Fix HIP/ROCm compilation some more

cp2k · Mar 12, 2022 · 66eea6d · 66eea6d
1 parent e563fa0
commit 66eea6d
Show file tree

Hide file tree

Showing 5 changed files with 21 additions and 16 deletions.
diff --git a/src/offload/offload_library.c b/src/offload/offload_library.c
@@ -16,11 +16,8 @@
 
 #if defined(__OFFLOAD_PROFILING)
 #if defined(__OFFLOAD_CUDA)
-#include <cuda.h>
-#include <cuda_runtime.h>
 #include <nvToolsExt.h>
 #elif defined(__OFFLOAD_HIP) && defined(__HIP_PLATFORM_AMD__)
-#include <hip/hip_runtime_api.h>
 #include <roctracer/roctx.h>
 #endif
 #endif

diff --git a/src/offload/offload_operations.h b/src/offload/offload_operations.h
@@ -18,7 +18,7 @@
 #if defined(__OFFLOAD_CUDA)
 #include <cuda_runtime.h>
 #elif defined(__OFFLOAD_HIP)
-#include <hip/hip_runtime_api.h>
+#include <hip/hip_runtime.h>
 #endif
 
 #ifdef __cplusplus

diff --git a/src/pw/gpu/pw_gpu_kernels.h b/src/pw/gpu/pw_gpu_kernels.h
@@ -16,29 +16,29 @@ extern "C" {
 
 /*******************************************************************************
  * \brief Launcher for pw_real_to_complex kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_real_to_complex(const double *din, double *zout,
                                    const int ngpts, offloadStream_t stream);
 
 /*******************************************************************************
  * \brief Launcher for pw_complex_to_real kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_complex_to_real(const double *zin, double *dout,
                                    const int ngpts, offloadStream_t stream);
 
 /*******************************************************************************
  * \brief Launcher for pw_gather_z kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_gather_z(double *pwcc, const double *c, const double scale,
                             const int ngpts, const int *ghatmap,
                             offloadStream_t stream);
 
 /*******************************************************************************
  * \brief Launcher for pw_scatter_z kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_scatter_z(double *c, const double *pwcc, const double scale,
                              const int ngpts, const int nmaps,

diff --git a/src/pw/gpu/pw_gpu_kernels_cuda.cu b/src/pw/gpu/pw_gpu_kernels_cuda.cu
@@ -28,10 +28,10 @@ __global__ void pw_real_to_complex(const double *din, double *zout,
 
 /*******************************************************************************
  * \brief Launcher for pw_real_to_complex kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_real_to_complex(const double *din, double *zout,
-                                   const int ngpts, cudaStream_t stream) {
+                                   const int ngpts, offloadStream_t stream) {
   const int threadsPerBlock = 1024;
   const int numBlocks = (ngpts + threadsPerBlock - 1) / threadsPerBlock;
   pw_real_to_complex<<<numBlocks, threadsPerBlock, 0, stream>>>(din, zout,
@@ -58,7 +58,7 @@ __global__ void pw_complex_to_real(const double *zin, double *dout,
 
 /*******************************************************************************
  * \brief Launcher for pw_complex_to_real kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_complex_to_real(const double *zin, double *dout,
                                    const int ngpts, offloadStream_t stream) {
@@ -83,7 +83,7 @@ __global__ void pw_gather_z(double *pwcc, const double *c, const double scale,
 
 /*******************************************************************************
  * \brief Launcher for pw_gather_z kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_gather_z(double *pwcc, const double *c, const double scale,
                             const int ngpts, const int *ghatmap,
@@ -114,7 +114,7 @@ __global__ void pw_scatter_z(double *c, const double *pwcc, const double scale,
 
 /*******************************************************************************
  * \brief Launcher for pw_scatter_z kernel.
- * \author Ole Schütt
+ * \author Ole Schuett
  ******************************************************************************/
 void pw_gpu_launch_scatter_z(double *c, const double *pwcc, const double scale,
                              const int ngpts, const int nmaps,

diff --git a/tools/toolchain/scripts/generate_arch_files.sh b/tools/toolchain/scripts/generate_arch_files.sh
@@ -161,30 +161,38 @@ if [ "${ENABLE_HIP}" = __TRUE__ ] && [ "${GPUVER}" != no ]; then
   add_lib_from_paths HIP_LDFLAGS "libhipfft.*" $LIB_PATHS
 
   PLATFORM_FLAGS=''
-  HIP_INCLUDES="-I${ROCM_PATH}/hip/include -I${ROCM_PATH}/hipblas/include -I${ROCM_PATH}/include"
+  HIP_INCLUDES="-I${ROCM_PATH}/include"
   case "${GPUVER}" in
     Mi50)
       check_lib -lamdhip64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libamdhip64.*" $LIB_PATHS
+      check_lib -lhipfft "hip"
+      add_lib_from_paths HIP_LDFLAGS "libhipfft.*" $LIB_PATHS
+      check_lib -lrocblas "hip"
+      add_lib_from_paths HIP_LDFLAGS "librocblas.*" $LIB_PATHS
       check_lib -lroctx64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libroctx64.*" $LIB_PATHS
       check_lib -lroctracer64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libroctracer64.*" $LIB_PATHS
       HIP_FLAGS+="-fPIE -D__HIP_PLATFORM_AMD__ -g --offload-arch=gfx906 -O3 --std=c++11 \$(DFLAGS)"
-      LIBS+=" IF_HIP(-lhipblas -lamdhip64 IF_DEBUG(-lroctx64 -lroctracer64|)|)"
+      LIBS+=" IF_HIP(-lamdhip64 -lhipfft -lhipblas -lrocblas IF_DEBUG(-lroctx64 -lroctracer64|)|)"
       PLATFORM_FLAGS='-D__HIP_PLATFORM_AMD__'
       DFLAGS+=' IF_HIP(-D__GRID_HIP -D__HIP_PLATFORM_AMD__ -D__PW_HIP -D__PW_GPU IF_DEBUG(-D__OFFLOAD_PROFILING|)|) -D__DBCSR_ACC'
       CXXFLAGS+=" -fopenmp -std=c++11"
       ;;
     Mi100)
       check_lib -lamdhip64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libamdhip64.*" $LIB_PATHS
+      check_lib -lhipfft "hip"
+      add_lib_from_paths HIP_LDFLAGS "libhipfft.*" $LIB_PATHS
+      check_lib -lrocblas "hip"
+      add_lib_from_paths HIP_LDFLAGS "librocblas.*" $LIB_PATHS
       check_lib -lroctx64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libroctx64.*" $LIB_PATHS
       check_lib -lroctracer64 "hip"
       add_lib_from_paths HIP_LDFLAGS "libroctracer64.*" $LIB_PATHS
       HIP_FLAGS+="-fPIE -D__HIP_PLATFORM_AMD__ -g --offload-arch=gfx908 -O3 --std=c++11 \$(DFLAGS)"
-      LIBS+=" IF_HIP( -lhipblas -lamdhip64 IF_DEBUG(-lroctx64 -lroctracer64|)|)"
+      LIBS+=" IF_HIP(-lamdhip64 -lhipfft -lhipblas -lrocblas IF_DEBUG(-lroctx64 -lroctracer64|)|)"
       PLATFORM_FLAGS='-D__HIP_PLATFORM_AMD__ '
       DFLAGS+=' IF_HIP(-D__GRID_HIP -D__HIP_PLATFORM_AMD__ -D__PW_HIP -D__PW_GPU IF_DEBUG(-D__OFFLOAD_PROFILING|)|) -D__DBCSR_ACC'
       CXXFLAGS+=" -fopenmp -std=c++11"