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 { 76 return cusolverDnDestroy(hipsolverhandle); 77 } 78 79 /* Alias hipsolverCreate to cusolverDnCreate */ 80 static inline hipsolverStatus_t hipsolverCreate(hipsolverHandle_t *hipsolverhandle) 81 { 82 return cusolverDnCreate(hipsolverhandle); 83 } 84 85 /* Alias hipsolverGetStream to cusolverDnGetStream */ 86 static inline hipsolverStatus_t hipsolverGetStream(hipsolverHandle_t handle, hipStream_t *stream) 87 { 88 return cusolverDnGetStream(handle, stream); 89 } 90 91 /* Alias hipsolverSetStream to cusolverDnSetStream */ 92 static inline hipsolverStatus_t hipsolverSetStream(hipsolverHandle_t handle, hipStream_t stream) 93 { 94 return cusolveDnSetStream(handle, stream); 95 } 96 #else /* __HIP_PLATFORM_HCC__ */ 97 typedef rocblas_handle hipsolverHandle_t; 98 typedef rocblas_status hipsolverStatus_t; 99 100 /* Alias hipsolverDestroy to rocblas_destroy_handle */ 101 static inline hipsolverStatus_t hipsolverDestroy(hipsolverHandle_t hipsolverhandle) 102 { 103 return rocblas_destroy_handle(hipsolverhandle); 104 } 105 106 /* Alias hipsolverCreate to rocblas_destroy_handle */ 107 static inline hipsolverStatus_t hipsolverCreate(hipsolverHandle_t *hipsolverhandle) 108 { 109 return rocblas_create_handle(hipsolverhandle); 110 } 111 112 // Alias hipsolverGetStream to rocblas_get_stream 113 static inline hipsolverStatus_t hipsolverGetStream(hipsolverHandle_t handle, hipStream_t *stream) 114 { 115 return rocblas_get_stream(handle, stream); 116 } 117 118 // Alias hipsolverSetStream to rocblas_set_stream 119 static inline hipsolverStatus_t hipsolverSetStream(hipsolverHandle_t handle, hipStream_t stream) 120 { 121 return rocblas_set_stream(handle, stream); 122 } 123 #endif // __HIP_PLATFORM_NVCC__ 124 125 // REMOVE ME 126 PETSC_EXTERN hipStream_t PetscDefaultHipStream; // The default stream used by PETSc 127 PETSC_EXTERN PetscErrorCode PetscHIPBLASGetHandle(hipblasHandle_t *); 128 PETSC_EXTERN PetscErrorCode PetscHIPSOLVERGetHandle(hipsolverHandle_t *); 129 130 #endif // PETSC_HAVE_HIP 131 132 // these can also be defined in petscdevice_cuda.h 133 #ifndef PETSC_DEVICE_DEFINED_DECLS_PRIVATE 134 #define PETSC_DEVICE_DEFINED_DECLS_PRIVATE 135 #if PetscDefined(USING_HCC) 136 #define PETSC_HOST_DECL __host__ 137 #define PETSC_DEVICE_DECL __device__ 138 #define PETSC_KERNEL_DECL __global__ 139 #define PETSC_SHAREDMEM_DECL __shared__ 140 #define PETSC_FORCEINLINE __forceinline__ 141 #define PETSC_CONSTMEM_DECL __constant__ 142 #else 143 #define PETSC_HOST_DECL 144 #define PETSC_DEVICE_DECL 145 #define PETSC_KERNEL_DECL 146 #define PETSC_SHAREDMEM_DECL 147 #define PETSC_FORCEINLINE inline 148 #define PETSC_CONSTMEM_DECL 149 #endif // PETSC_USING_NVCC 150 151 #define PETSC_HOSTDEVICE_DECL PETSC_HOST_DECL PETSC_DEVICE_DECL 152 #define PETSC_DEVICE_INLINE_DECL PETSC_DEVICE_DECL PETSC_FORCEINLINE 153 #define PETSC_HOSTDEVICE_INLINE_DECL PETSC_HOSTDEVICE_DECL PETSC_FORCEINLINE 154 #endif // PETSC_DEVICE_DEFINED_DECLS_PRIVATE 155 156 #endif // PETSCDEVICE_HIP_H 157