Actual source code: cupmblasinterface.hpp

  1: #pragma once

  3: #include <petsc/private/cupminterface.hpp>
  4: #include <petsc/private/petscadvancedmacros.h>

  6: #include <limits> // std::numeric_limits

  8: namespace Petsc
  9: {

 11: namespace device
 12: {

 14: namespace cupm
 15: {

 17: namespace impl
 18: {

 20: #define PetscCallCUPMBLAS_(__abort_fn__, __comm__, ...) \
 21:   do { \
 22:     PetscStackUpdateLine; \
 23:     const cupmBlasError_t cberr_p_ = __VA_ARGS__; \
 24:     if (PetscUnlikely(cberr_p_ != CUPMBLAS_STATUS_SUCCESS)) { \
 25:       if (((cberr_p_ == CUPMBLAS_STATUS_NOT_INITIALIZED) || (cberr_p_ == CUPMBLAS_STATUS_ALLOC_FAILED)) && PetscDeviceInitialized(PETSC_DEVICE_CUPM())) { \
 26:         __abort_fn__(__comm__, PETSC_ERR_GPU_RESOURCE, \
 27:                      "%s error %d (%s). Reports not initialized or alloc failed; " \
 28:                      "this indicates the GPU may have run out resources", \
 29:                      cupmBlasName(), static_cast<PetscErrorCode>(cberr_p_), cupmBlasGetErrorName(cberr_p_)); \
 30:       } \
 31:       __abort_fn__(__comm__, PETSC_ERR_GPU, "%s error %d (%s)", cupmBlasName(), static_cast<PetscErrorCode>(cberr_p_), cupmBlasGetErrorName(cberr_p_)); \
 32:     } \
 33:   } while (0)

 35: #define PetscCallCUPMBLAS(...)             PetscCallCUPMBLAS_(SETERRQ, PETSC_COMM_SELF, __VA_ARGS__)
 36: #define PetscCallCUPMBLASAbort(comm_, ...) PetscCallCUPMBLAS_(SETERRABORT, comm_, __VA_ARGS__)

 38: // given cupmBlasaxpy() then
 39: // T = PETSC_CUPBLAS_FP_TYPE
 40: // given cupmBlasnrm2() then
 41: // T = PETSC_CUPMBLAS_FP_INPUT_TYPE
 42: // u = PETSC_CUPMBLAS_FP_RETURN_TYPE
 43: #if PetscDefined(USE_COMPLEX)
 44:   #if PetscDefined(USE_REAL_SINGLE)
 45:     #define PETSC_CUPMBLAS_FP_TYPE_U       C
 46:     #define PETSC_CUPMBLAS_FP_TYPE_L       c
 47:     #define PETSC_CUPMBLAS_FP_INPUT_TYPE_U S
 48:     #define PETSC_CUPMBLAS_FP_INPUT_TYPE_L s
 49:   #elif PetscDefined(USE_REAL_DOUBLE)
 50:     #define PETSC_CUPMBLAS_FP_TYPE_U       Z
 51:     #define PETSC_CUPMBLAS_FP_TYPE_L       z
 52:     #define PETSC_CUPMBLAS_FP_INPUT_TYPE_U D
 53:     #define PETSC_CUPMBLAS_FP_INPUT_TYPE_L d
 54:   #endif
 55:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE_U PETSC_CUPMBLAS_FP_TYPE_U
 56:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE_L PETSC_CUPMBLAS_FP_TYPE_L
 57: #else
 58:   #if PetscDefined(USE_REAL_SINGLE)
 59:     #define PETSC_CUPMBLAS_FP_TYPE_U S
 60:     #define PETSC_CUPMBLAS_FP_TYPE_L s
 61:   #elif PetscDefined(USE_REAL_DOUBLE)
 62:     #define PETSC_CUPMBLAS_FP_TYPE_U D
 63:     #define PETSC_CUPMBLAS_FP_TYPE_L d
 64:   #endif
 65:   #define PETSC_CUPMBLAS_FP_INPUT_TYPE_U PETSC_CUPMBLAS_FP_TYPE_U
 66:   #define PETSC_CUPMBLAS_FP_INPUT_TYPE_L PETSC_CUPMBLAS_FP_TYPE_L
 67:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE_U
 68:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE_L
 69: #endif // USE_COMPLEX

 71: #if !defined(PETSC_CUPMBLAS_FP_TYPE_U) && !PetscDefined(USE_REAL___FLOAT128)
 72:   #error "Unsupported floating-point type for CUDA/HIP BLAS"
 73: #endif

 75: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_MODIFIED() - Helper macro to build a "modified"
 76: // blas function whose return type does not match the input type
 77: //
 78: // input param:
 79: // func - base suffix of the blas function, e.g. nrm2
 80: //
 81: // notes:
 82: // requires PETSC_CUPMBLAS_FP_INPUT_TYPE to be defined as the blas floating point input type
 83: // letter ("S" for real/complex single, "D" for real/complex double).
 84: //
 85: // requires PETSC_CUPMBLAS_FP_RETURN_TYPE to be defined as the blas floating point output type
 86: // letter ("c" for complex single, "z" for complex double and  for real
 87: // single/double).
 88: //
 89: // In their infinite wisdom nvidia/amd have made the upper-case vs lower-case scheme
 90: // infuriatingly inconsistent...
 91: //
 92: // example usage:
 93: // #define PETSC_CUPMBLAS_FP_INPUT_TYPE  S
 94: // #define PETSC_CUPMBLAS_FP_RETURN_TYPE
 95: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_MODIFIED(nrm2) -> Snrm2
 96: //
 97: // #define PETSC_CUPMBLAS_FP_INPUT_TYPE  D
 98: // #define PETSC_CUPMBLAS_FP_RETURN_TYPE z
 99: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_MODIFIED(nrm2) -> Dznrm2
100: #define PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_MODIFIED(func) PetscConcat(PetscConcat(PETSC_CUPMBLAS_FP_INPUT_TYPE, PETSC_CUPMBLAS_FP_RETURN_TYPE), func)

102: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_IFPTYPE() - Helper macro to build Iamax and Iamin
103: // because they are both extra special
104: //
105: // input param:
106: // func - base suffix of the blas function, either amax or amin
107: //
108: // notes:
109: // The macro name literally stands for "I" ## "floating point type" because shockingly enough,
110: // that's what it does.
111: //
112: // requires PETSC_CUPMBLAS_FP_TYPE_L to be defined as the lower-case blas floating point input type
113: // letter ("s" for complex single, "z" for complex double, "s" for real single, and "d" for
114: // real double).
115: //
116: // example usage:
117: // #define PETSC_CUPMBLAS_FP_TYPE_L s
118: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_IFPTYPE(amax) -> Isamax
119: //
120: // #define PETSC_CUPMBLAS_FP_TYPE_L z
121: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_IFPTYPE(amin) -> Izamin
122: #define PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_IFPTYPE(func) PetscConcat(I, PetscConcat(PETSC_CUPMBLAS_FP_TYPE_L, func))

124: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_STANDARD() - Helper macro to build a "standard"
125: // blas function name
126: //
127: // input param:
128: // func - base suffix of the blas function, e.g. axpy, scal
129: //
130: // notes:
131: // requires PETSC_CUPMBLAS_FP_TYPE to be defined as the blas floating-point letter ("C" for
132: // complex single, "Z" for complex double, "S" for real single, "D" for real double).
133: //
134: // example usage:
135: // #define PETSC_CUPMBLAS_FP_TYPE S
136: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_STANDARD(axpy) -> Saxpy
137: //
138: // #define PETSC_CUPMBLAS_FP_TYPE Z
139: // PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_STANDARD(axpy) -> Zaxpy
140: #define PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_STANDARD(func) PetscConcat(PETSC_CUPMBLAS_FP_TYPE, func)

142: // PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT() - In case CUDA/HIP don't agree with our suffix
143: // one can provide both here
144: //
145: // input params:
146: // MACRO_SUFFIX - suffix to one of the above blas function builder macros, e.g. STANDARD or
147: // IFPTYPE
148: // our_suffix   - the suffix of the alias function
149: // their_suffix - the suffix of the function being aliased
150: //
151: // notes:
152: // requires PETSC_CUPMBLAS_PREFIX to be defined as the specific CUDA/HIP blas function
153: // prefix. requires any other specific definitions required by the specific builder macro to
154: // also be defined. See PETSC_CUPM_ALIAS_FUNCTION_EXACT() for the exact expansion of the
155: // function alias.
156: //
157: // example usage:
158: // #define PETSC_CUPMBLAS_PREFIX  cublas
159: // #define PETSC_CUPMBLAS_FP_TYPE C
160: // PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD,dot,dotc) ->
161: // template 
162: // static constexpr auto cupmBlasXdot(T&&... args) *noexcept and returntype detection*
163: // {
164: //   return cublasCdotc(std::forward(args)...);
165: // }
166: #define PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(MACRO_SUFFIX, our_suffix, their_suffix) \
167:   PETSC_CUPM_ALIAS_FUNCTION(PetscConcat(cupmBlasX, our_suffix), PetscConcat(PETSC_CUPMBLAS_PREFIX, PetscConcat(PETSC_CUPMBLAS_BUILD_BLAS_FUNCTION_ALIAS_, MACRO_SUFFIX)(their_suffix)))

169: // PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION() - Alias a CUDA/HIP blas function
170: //
171: // input params:
172: // MACRO_SUFFIX - suffix to one of the above blas function builder macros, e.g. STANDARD or
173: // IFPTYPE
174: // suffix       - the common suffix between CUDA and HIP of the alias function
175: //
176: // notes:
177: // see PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(), this macro just calls that one with "suffix" as
178: // "our_prefix" and "their_prefix"
179: #define PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(MACRO_SUFFIX, suffix) PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(MACRO_SUFFIX, suffix, suffix)

181: // PETSC_CUPMBLAS_ALIAS_FUNCTION() - Alias a CUDA/HIP library function
182: //
183: // input params:
184: // suffix - the common suffix between CUDA and HIP of the alias function
185: //
186: // notes:
187: // requires PETSC_CUPMBLAS_PREFIX to be defined as the specific CUDA/HIP blas library
188: // prefix. see PETSC_CUPMM_ALIAS_FUNCTION_EXACT() for the precise expansion of this macro.
189: //
190: // example usage:
191: // #define PETSC_CUPMBLAS_PREFIX hipblas
192: // PETSC_CUPMBLAS_ALIAS_FUNCTION(Create) ->
193: // template 
194: // static constexpr auto cupmBlasCreate(T&&... args) *noexcept and returntype detection*
195: // {
196: //   return hipblasCreate(std::forward(args)...);
197: // }
198: #define PETSC_CUPMBLAS_ALIAS_FUNCTION(suffix) PETSC_CUPM_ALIAS_FUNCTION(PetscConcat(cupmBlas, suffix), PetscConcat(PETSC_CUPMBLAS_PREFIX, suffix))

200: template <DeviceType>
201: struct BlasInterfaceImpl;

203: // Exists because HIP (for whatever godforsaken reason) has elected to define both their
204: // hipBlasHandle_t and hipSolverHandle_t as void *. So we cannot disambiguate them for overload
205: // resolution and hence need to wrap their types int this mess.
206: template <typename T, std::size_t I>
207: class cupmBlasHandleWrapper {
208: public:
209:   constexpr cupmBlasHandleWrapper() noexcept = default;
210:   constexpr cupmBlasHandleWrapper(T h) noexcept : handle_{std::move(h)} { static_assert(std::is_standard_layout<cupmBlasHandleWrapper<T, I>>::value, ""); }

212:   cupmBlasHandleWrapper &operator=(std::nullptr_t) noexcept
213:   {
214:     handle_ = nullptr;
215:     return *this;
216:   }

218:   operator T() const { return handle_; }

220:   const T *ptr_to() const { return &handle_; }
221:   T       *ptr_to() { return &handle_; }

223: private:
224:   T handle_{};
225: };

227: #if PetscDefined(HAVE_CUDA)
228:   #define PETSC_CUPMBLAS_PREFIX         cublas
229:   #define PETSC_CUPMBLAS_PREFIX_U       CUBLAS
230:   #define PETSC_CUPMBLAS_FP_TYPE        PETSC_CUPMBLAS_FP_TYPE_U
231:   #define PETSC_CUPMBLAS_FP_INPUT_TYPE  PETSC_CUPMBLAS_FP_INPUT_TYPE_U
232:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE PETSC_CUPMBLAS_FP_RETURN_TYPE_L
233: template <>
234: struct PETSC_SINGLE_LIBRARY_VISIBILITY_INTERNAL BlasInterfaceImpl<DeviceType::CUDA> : Interface<DeviceType::CUDA> {
235:   // typedefs
236:   using cupmBlasHandle_t      = cupmBlasHandleWrapper<cublasHandle_t, 0>;
237:   using cupmBlasError_t       = cublasStatus_t;
238:   using cupmBlasInt_t         = int;
239:   using cupmBlasPointerMode_t = cublasPointerMode_t;

241:   // values
242:   static const auto CUPMBLAS_STATUS_SUCCESS         = CUBLAS_STATUS_SUCCESS;
243:   static const auto CUPMBLAS_STATUS_NOT_INITIALIZED = CUBLAS_STATUS_NOT_INITIALIZED;
244:   static const auto CUPMBLAS_STATUS_ALLOC_FAILED    = CUBLAS_STATUS_ALLOC_FAILED;
245:   static const auto CUPMBLAS_POINTER_MODE_HOST      = CUBLAS_POINTER_MODE_HOST;
246:   static const auto CUPMBLAS_POINTER_MODE_DEVICE    = CUBLAS_POINTER_MODE_DEVICE;
247:   static const auto CUPMBLAS_OP_T                   = CUBLAS_OP_T;
248:   static const auto CUPMBLAS_OP_N                   = CUBLAS_OP_N;
249:   static const auto CUPMBLAS_OP_C                   = CUBLAS_OP_C;
250:   static const auto CUPMBLAS_FILL_MODE_LOWER        = CUBLAS_FILL_MODE_LOWER;
251:   static const auto CUPMBLAS_FILL_MODE_UPPER        = CUBLAS_FILL_MODE_UPPER;
252:   static const auto CUPMBLAS_SIDE_LEFT              = CUBLAS_SIDE_LEFT;
253:   static const auto CUPMBLAS_DIAG_NON_UNIT          = CUBLAS_DIAG_NON_UNIT;

255:   // utility functions
256:   PETSC_CUPMBLAS_ALIAS_FUNCTION(Create)
257:   PETSC_CUPMBLAS_ALIAS_FUNCTION(Destroy)
258:   PETSC_CUPMBLAS_ALIAS_FUNCTION(GetStream)
259:   PETSC_CUPMBLAS_ALIAS_FUNCTION(SetStream)
260:   PETSC_CUPMBLAS_ALIAS_FUNCTION(GetPointerMode)
261:   PETSC_CUPMBLAS_ALIAS_FUNCTION(SetPointerMode)

263:   // level 1 BLAS
264:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, axpy)
265:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, copy)
266:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, scal)
267:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, dot, PetscIfPetscDefined(USE_COMPLEX, dotc, dot))
268:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, dotu, PetscIfPetscDefined(USE_COMPLEX, dotu, dot))
269:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, swap)
270:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(MODIFIED, nrm2)
271:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(IFPTYPE, amax)
272:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(MODIFIED, asum)

274:   // level 2 BLAS
275:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gemv)
276:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trmv)
277:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trsv)
278:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gbmv)
279:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, tbmv)
280:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, tbsv)
281:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, hemv, PetscIfPetscDefined(USE_COMPLEX, hemv, symv))

283:   // level 3 BLAS
284:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gemm)
285:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trsm)

287:   // BLAS extensions
288:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, geam)

290:   PETSC_NODISCARD static const char *cupmBlasGetErrorName(cupmBlasError_t status) noexcept { return PetscCUBLASGetErrorName(status); }
291: };
292:   #undef PETSC_CUPMBLAS_PREFIX
293:   #undef PETSC_CUPMBLAS_PREFIX_U
294:   #undef PETSC_CUPMBLAS_FP_TYPE
295:   #undef PETSC_CUPMBLAS_FP_INPUT_TYPE
296:   #undef PETSC_CUPMBLAS_FP_RETURN_TYPE
297: #endif // PetscDefined(HAVE_CUDA)

299: #if PetscDefined(HAVE_HIP)
300:   #define PETSC_CUPMBLAS_PREFIX         hipblas
301:   #define PETSC_CUPMBLAS_PREFIX_U       HIPBLAS
302:   #define PETSC_CUPMBLAS_FP_TYPE        PETSC_CUPMBLAS_FP_TYPE_U
303:   #define PETSC_CUPMBLAS_FP_INPUT_TYPE  PETSC_CUPMBLAS_FP_INPUT_TYPE_U
304:   #define PETSC_CUPMBLAS_FP_RETURN_TYPE PETSC_CUPMBLAS_FP_RETURN_TYPE_L
305: template <>
306: struct PETSC_SINGLE_LIBRARY_VISIBILITY_INTERNAL BlasInterfaceImpl<DeviceType::HIP> : Interface<DeviceType::HIP> {
307:   // typedefs
308:   using cupmBlasHandle_t      = cupmBlasHandleWrapper<hipblasHandle_t, 0>;
309:   using cupmBlasError_t       = hipblasStatus_t;
310:   using cupmBlasInt_t         = int; // rocblas will have its own
311:   using cupmBlasPointerMode_t = hipblasPointerMode_t;

313:   // values
314:   static const auto CUPMBLAS_STATUS_SUCCESS         = HIPBLAS_STATUS_SUCCESS;
315:   static const auto CUPMBLAS_STATUS_NOT_INITIALIZED = HIPBLAS_STATUS_NOT_INITIALIZED;
316:   static const auto CUPMBLAS_STATUS_ALLOC_FAILED    = HIPBLAS_STATUS_ALLOC_FAILED;
317:   static const auto CUPMBLAS_POINTER_MODE_HOST      = HIPBLAS_POINTER_MODE_HOST;
318:   static const auto CUPMBLAS_POINTER_MODE_DEVICE    = HIPBLAS_POINTER_MODE_DEVICE;
319:   static const auto CUPMBLAS_OP_T                   = HIPBLAS_OP_T;
320:   static const auto CUPMBLAS_OP_N                   = HIPBLAS_OP_N;
321:   static const auto CUPMBLAS_OP_C                   = HIPBLAS_OP_C;
322:   static const auto CUPMBLAS_FILL_MODE_LOWER        = HIPBLAS_FILL_MODE_LOWER;
323:   static const auto CUPMBLAS_FILL_MODE_UPPER        = HIPBLAS_FILL_MODE_UPPER;
324:   static const auto CUPMBLAS_SIDE_LEFT              = HIPBLAS_SIDE_LEFT;
325:   static const auto CUPMBLAS_DIAG_NON_UNIT          = HIPBLAS_DIAG_NON_UNIT;

327:   // utility functions
328:   PETSC_CUPMBLAS_ALIAS_FUNCTION(Create)
329:   PETSC_CUPMBLAS_ALIAS_FUNCTION(Destroy)
330:   PETSC_CUPMBLAS_ALIAS_FUNCTION(GetStream)
331:   PETSC_CUPMBLAS_ALIAS_FUNCTION(SetStream)
332:   PETSC_CUPMBLAS_ALIAS_FUNCTION(GetPointerMode)
333:   PETSC_CUPMBLAS_ALIAS_FUNCTION(SetPointerMode)

335:   // level 1 BLAS
336:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, axpy)
337:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, copy)
338:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, scal)
339:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, dot, PetscIfPetscDefined(USE_COMPLEX, dotc, dot))
340:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, dotu, PetscIfPetscDefined(USE_COMPLEX, dotu, dot))
341:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, swap)
342:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(MODIFIED, nrm2)
343:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(IFPTYPE, amax)
344:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(MODIFIED, asum)

346:   // level 2 BLAS
347:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gemv)
348:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trmv)
349:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trsv)
350:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gbmv)
351:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, tbmv)
352:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, tbsv)
353:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION_EXACT(STANDARD, hemv, PetscIfPetscDefined(USE_COMPLEX, hemv, symv))

355:   // level 3 BLAS
356:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, gemm)
357:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, trsm)

359:   // BLAS extensions
360:   PETSC_CUPMBLAS_ALIAS_BLAS_FUNCTION(STANDARD, geam)

362:   PETSC_NODISCARD static const char *cupmBlasGetErrorName(cupmBlasError_t status) noexcept { return PetscHIPBLASGetErrorName(status); }
363: };
364:   #undef PETSC_CUPMBLAS_PREFIX
365:   #undef PETSC_CUPMBLAS_PREFIX_U
366:   #undef PETSC_CUPMBLAS_FP_TYPE
367:   #undef PETSC_CUPMBLAS_FP_INPUT_TYPE
368:   #undef PETSC_CUPMBLAS_FP_RETURN_TYPE
369: #endif // PetscDefined(HAVE_HIP)

371: #define PETSC_CUPMBLAS_IMPL_CLASS_HEADER(T) \
372:   PETSC_CUPM_INHERIT_INTERFACE_TYPEDEFS_USING(T); \
373:   /* introspection */ \
374:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasGetErrorName; \
375:   /* types */ \
376:   using cupmBlasHandle_t      = typename ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasHandle_t; \
377:   using cupmBlasError_t       = typename ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasError_t; \
378:   using cupmBlasInt_t         = typename ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasInt_t; \
379:   using cupmBlasPointerMode_t = typename ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasPointerMode_t; \
380:   /* values */ \
381:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_STATUS_SUCCESS; \
382:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_STATUS_NOT_INITIALIZED; \
383:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_STATUS_ALLOC_FAILED; \
384:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_POINTER_MODE_HOST; \
385:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_POINTER_MODE_DEVICE; \
386:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_OP_T; \
387:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_OP_N; \
388:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_OP_C; \
389:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_FILL_MODE_LOWER; \
390:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_FILL_MODE_UPPER; \
391:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_SIDE_LEFT; \
392:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::CUPMBLAS_DIAG_NON_UNIT; \
393:   /* utility functions */ \
394:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasCreate; \
395:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasDestroy; \
396:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasGetStream; \
397:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasSetStream; \
398:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasGetPointerMode; \
399:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasSetPointerMode; \
400:   /* level 1 BLAS */ \
401:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXaxpy; \
402:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXcopy; \
403:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXscal; \
404:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXdot; \
405:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXdotu; \
406:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXswap; \
407:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXnrm2; \
408:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXamax; \
409:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXasum; \
410:   /* level 2 BLAS */ \
411:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXgemv; \
412:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXtrmv; \
413:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXtrsv; \
414:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXgbmv; \
415:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXtbmv; \
416:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXtbsv; \
417:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXhemv; \
418:   /* level 3 BLAS */ \
419:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXgemm; \
420:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXtrsm; \
421:   /* BLAS extensions */ \
422:   using ::Petsc::device::cupm::impl::BlasInterfaceImpl<T>::cupmBlasXgeam

424: // The actual interface class
425: template <DeviceType T>
426: struct BlasInterface : BlasInterfaceImpl<T> {
427:   PETSC_CUPMBLAS_IMPL_CLASS_HEADER(T);

429:   PETSC_NODISCARD static constexpr const char *cupmBlasName() noexcept { return T == DeviceType::CUDA ? "cuBLAS" : "hipBLAS"; }

431:   static PetscErrorCode PetscCUPMBlasSetPointerModeFromPointer(cupmBlasHandle_t handle, const void *ptr) noexcept
432:   {
433:     auto mtype = PETSC_MEMTYPE_HOST;

435:     PetscFunctionBegin;
436:     PetscCall(PetscCUPMGetMemType(ptr, &mtype));
437:     PetscCallCUPMBLAS(cupmBlasSetPointerMode(handle, PetscMemTypeDevice(mtype) ? CUPMBLAS_POINTER_MODE_DEVICE : CUPMBLAS_POINTER_MODE_HOST));
438:     PetscFunctionReturn(PETSC_SUCCESS);
439:   }

441:   static PetscErrorCode checkCupmBlasIntCast(PetscInt x) noexcept
442:   {
443:     PetscFunctionBegin;
444:     PetscCheck((std::is_same<PetscInt, cupmBlasInt_t>::value) || (x <= std::numeric_limits<cupmBlasInt_t>::max()), PETSC_COMM_SELF, PETSC_ERR_ARG_OUTOFRANGE, "%" PetscInt_FMT " is too big for %s, which may be restricted to 32-bit integers", x, cupmBlasName());
445:     PetscCheck(x >= 0, PETSC_COMM_SELF, PETSC_ERR_ARG_OUTOFRANGE, "Passing negative integer (%" PetscInt_FMT ") to %s routine", x, cupmBlasName());
446:     PetscFunctionReturn(PETSC_SUCCESS);
447:   }

449:   static PetscErrorCode PetscCUPMBlasIntCast(PetscInt x, cupmBlasInt_t *y) noexcept
450:   {
451:     PetscFunctionBegin;
452:     *y = static_cast<cupmBlasInt_t>(x);
453:     PetscCall(checkCupmBlasIntCast(x));
454:     PetscFunctionReturn(PETSC_SUCCESS);
455:   }

457:   class CUPMBlasPointerModeGuard {
458:   public:
459:     CUPMBlasPointerModeGuard(const cupmBlasHandle_t &handle, cupmBlasPointerMode_t mode) noexcept : handle_{handle}
460:     {
461:       PetscFunctionBegin;
462:       PetscCallCUPMBLASAbort(PETSC_COMM_SELF, cupmBlasGetPointerMode(handle, &this->old_));
463:       if (this->old_ == mode) {
464:         this->set_ = false;
465:       } else {
466:         this->set_ = true;
467:         PetscCallCUPMBLASAbort(PETSC_COMM_SELF, cupmBlasSetPointerMode(handle, mode));
468:       }
469:       PetscFunctionReturnVoid();
470:     }

472:     CUPMBlasPointerModeGuard(const cupmBlasHandle_t &handle, PetscMemType mtype) noexcept : CUPMBlasPointerModeGuard{handle, PetscMemTypeDevice(mtype) ? CUPMBLAS_POINTER_MODE_DEVICE : CUPMBLAS_POINTER_MODE_HOST} { }

474:     ~CUPMBlasPointerModeGuard() noexcept
475:     {
476:       PetscFunctionBegin;
477:       if (this->set_) PetscCallCUPMBLASAbort(PETSC_COMM_SELF, cupmBlasSetPointerMode(this->handle_, this->old_));
478:       PetscFunctionReturnVoid();
479:     }

481:   private:
482:     cupmBlasHandle_t      handle_;
483:     cupmBlasPointerMode_t old_;
484:     bool                  set_;
485:   };
486: };

488: #define PETSC_CUPMBLAS_INHERIT_INTERFACE_TYPEDEFS_USING(T) \
489:   PETSC_CUPMBLAS_IMPL_CLASS_HEADER(T); \
490:   using ::Petsc::device::cupm::impl::BlasInterface<T>::cupmBlasName; \
491:   using ::Petsc::device::cupm::impl::BlasInterface<T>::PetscCUPMBlasSetPointerModeFromPointer; \
492:   using ::Petsc::device::cupm::impl::BlasInterface<T>::checkCupmBlasIntCast; \
493:   using ::Petsc::device::cupm::impl::BlasInterface<T>::PetscCUPMBlasIntCast; \
494:   using CUPMBlasPointerModeGuard = typename ::Petsc::device::cupm::impl::BlasInterface<T>::CUPMBlasPointerModeGuard

496: #if PetscDefined(HAVE_CUDA)
497: extern template struct PETSC_SINGLE_LIBRARY_VISIBILITY_INTERNAL BlasInterface<DeviceType::CUDA>;
498: #endif

500: #if PetscDefined(HAVE_HIP)
501: extern template struct PETSC_SINGLE_LIBRARY_VISIBILITY_INTERNAL BlasInterface<DeviceType::HIP>;
502: #endif

504: } // namespace impl

506: } // namespace cupm

508: } // namespace device

510: } // namespace Petsc