1 #ifndef PETSCDEVICE_HIP_H 2 #define PETSCDEVICE_HIP_H 3 4 #include <petscdevice.h> 5 #include <petscpkg_version.h> 6 7 #if defined(__HCC__) || (defined(__clang__) && defined(__HIP__)) 8 #define PETSC_USING_HCC 1 9 #endif 10 11 #if PetscDefined(HAVE_HIP) 12 #include <hip/hip_runtime.h> 13 14 #if PETSC_PKG_HIP_VERSION_GE(5, 2, 0) 15 #include <hipblas/hipblas.h> 16 #else 17 #include <hipblas.h> 18 #endif 19 20 #if defined(__HIP_PLATFORM_NVCC__) 21 #include <cusolverDn.h> 22 #else // __HIP_PLATFORM_HCC__ 23 #if PETSC_PKG_HIP_VERSION_GE(5, 2, 0) 24 #include <rocsolver/rocsolver.h> 25 #else 26 #include <rocsolver.h> 27 #endif 28 #endif // __HIP_PLATFORM_NVCC__ 29 #include <hip/hip_complex.h> // for hipComplex, hipDoubleComplex 30 31 // REMOVE ME 32 #define WaitForHIP() hipDeviceSynchronize() 33 34 /* hipBLAS does not have hipblasGetErrorName(). We create one on our own. */ 35 PETSC_EXTERN const char *PetscHIPBLASGetErrorName(hipblasStatus_t); /* PETSC_EXTERN since it is exposed by the CHKERRHIPBLAS macro */ 36 37 #define PetscCallHIP(...) \ 38 do { \ 39 const hipError_t _p_hip_err__ = __VA_ARGS__; \ 40 if (PetscUnlikely(_p_hip_err__ != hipSuccess)) { \ 41 const char *name = hipGetErrorName(_p_hip_err__); \ 42 const char *descr = hipGetErrorString(_p_hip_err__); \ 43 SETERRQ(PETSC_COMM_SELF, PETSC_ERR_GPU, "hip error %d (%s) : %s", (PetscErrorCode)_p_hip_err__, name, descr); \ 44 } \ 45 } while (0) 46 #define CHKERRHIP(...) PetscCallHIP(__VA_ARGS__) 47 48 #define PetscCallHIPBLAS(...) \ 49 do { \ 50 const hipblasStatus_t _p_hipblas_stat__ = __VA_ARGS__; \ 51 if (PetscUnlikely(_p_hipblas_stat__ != HIPBLAS_STATUS_SUCCESS)) { \ 52 const char *name = PetscHIPBLASGetErrorName(_p_hipblas_stat__); \ 53 SETERRQ(PETSC_COMM_SELF, PETSC_ERR_GPU, "hipBLAS error %d (%s)", (PetscErrorCode)_p_hipblas_stat__, name); \ 54 } \ 55 } while (0) 56 #define CHKERRHIPBLAS(...) PetscCallHIPBLAS(__VA_ARGS__) 57 58 /* TODO: SEK: Need to figure out the hipsolver issues */ 59 #define PetscCallHIPSOLVER(...) \ 60 do { \ 61 const hipsolverStatus_t _p_hipsolver_stat__ = __VA_ARGS__; \ 62 PetscCheck(!_p_hipsolver_stat__, PETSC_COMM_SELF, PETSC_ERR_GPU, "HIPSOLVER error %d", (PetscErrorCode)_p_hipsolver_stat__); \ 63 } while (0) 64 #define CHKERRHIPSOLVER(...) PetscCallHIPSOLVER(__VA_ARGS__) 65 66 /* hipSolver does not exist yet so we work around it 67 rocSOLVER users rocBLAS for the handle 68 * */ 69 #if defined(__HIP_PLATFORM_NVCC__) 70 typedef cusolverDnHandle_t hipsolverHandle_t; 71 typedef cusolverStatus_t hipsolverStatus_t; 72 73 /* Alias hipsolverDestroy to cusolverDnDestroy */ 74 static inline hipsolverStatus_t hipsolverDestroy(hipsolverHandle_t *hipsolverhandle) { 75 return cusolverDnDestroy(hipsolverhandle); 76 } 77 78 /* Alias hipsolverCreate to cusolverDnCreate */ 79 static inline hipsolverStatus_t hipsolverCreate(hipsolverHandle_t *hipsolverhandle) { 80 return cusolverDnCreate(hipsolverhandle); 81 } 82 83 /* Alias hipsolverGetStream to cusolverDnGetStream */ 84 static inline hipsolverStatus_t hipsolverGetStream(hipsolverHandle_t handle, hipStream_t *stream) { 85 return cusolverDnGetStream(handle, stream); 86 } 87 88 /* Alias hipsolverSetStream to cusolverDnSetStream */ 89 static inline hipsolverStatus_t hipsolverSetStream(hipsolverHandle_t handle, hipStream_t stream) { 90 return cusolveDnSetStream(handle, stream); 91 } 92 #else /* __HIP_PLATFORM_HCC__ */ 93 typedef rocblas_handle hipsolverHandle_t; 94 typedef rocblas_status hipsolverStatus_t; 95 96 /* Alias hipsolverDestroy to rocblas_destroy_handle */ 97 static inline hipsolverStatus_t hipsolverDestroy(hipsolverHandle_t hipsolverhandle) { 98 return rocblas_destroy_handle(hipsolverhandle); 99 } 100 101 /* Alias hipsolverCreate to rocblas_destroy_handle */ 102 static inline hipsolverStatus_t hipsolverCreate(hipsolverHandle_t *hipsolverhandle) { 103 return rocblas_create_handle(hipsolverhandle); 104 } 105 106 // Alias hipsolverGetStream to rocblas_get_stream 107 static inline hipsolverStatus_t hipsolverGetStream(hipsolverHandle_t handle, hipStream_t *stream) { 108 return rocblas_get_stream(handle, stream); 109 } 110 111 // Alias hipsolverSetStream to rocblas_set_stream 112 static inline hipsolverStatus_t hipsolverSetStream(hipsolverHandle_t handle, hipStream_t stream) { 113 return rocblas_set_stream(handle, stream); 114 } 115 #endif // __HIP_PLATFORM_NVCC__ 116 117 // REMOVE ME 118 PETSC_EXTERN hipStream_t PetscDefaultHipStream; // The default stream used by PETSc 119 PETSC_EXTERN PetscErrorCode PetscHIPBLASGetHandle(hipblasHandle_t *); 120 PETSC_EXTERN PetscErrorCode PetscHIPSOLVERGetHandle(hipsolverHandle_t *); 121 122 #endif // PETSC_HAVE_HIP 123 124 // these can also be defined in petscdevice_cuda.h 125 #ifndef PETSC_DEVICE_DEFINED_DECLS_PRIVATE 126 #define PETSC_DEVICE_DEFINED_DECLS_PRIVATE 127 #if PetscDefined(USING_HCC) 128 #define PETSC_HOST_DECL __host__ 129 #define PETSC_DEVICE_DECL __device__ 130 #define PETSC_KERNEL_DECL __global__ 131 #define PETSC_SHAREDMEM_DECL __shared__ 132 #define PETSC_FORCEINLINE __forceinline__ 133 #define PETSC_CONSTMEM_DECL __constant__ 134 #else 135 #define PETSC_HOST_DECL 136 #define PETSC_DEVICE_DECL 137 #define PETSC_KERNEL_DECL 138 #define PETSC_SHAREDMEM_DECL 139 #define PETSC_FORCEINLINE inline 140 #define PETSC_CONSTMEM_DECL 141 #endif // PETSC_USING_NVCC 142 143 #define PETSC_HOSTDEVICE_DECL PETSC_HOST_DECL PETSC_DEVICE_DECL 144 #define PETSC_DEVICE_INLINE_DECL PETSC_DEVICE_DECL PETSC_FORCEINLINE 145 #define PETSC_HOSTDEVICE_INLINE_DECL PETSC_HOSTDEVICE_DECL PETSC_FORCEINLINE 146 #endif // PETSC_DEVICE_DEFINED_DECLS_PRIVATE 147 148 #endif // PETSCDEVICE_HIP_H 149