refactor acc with a common utils

cp2k · Jan 25, 2024 · a8bad85 · a8bad85
1 parent 79cc06e
commit a8bad85
Show file tree

Hide file tree

Showing 6 changed files with 43 additions and 12 deletions.
diff --git a/src/CMakeLists.txt b/src/CMakeLists.txt
@@ -114,6 +114,7 @@ set(DBCSR_HIP_AND_CUDA_SRCS
     acc/cuda_hip/acc_dev.cpp
     acc/cuda_hip/acc_error.cpp
     acc/cuda_hip/acc_event.cpp
+    acc/cuda_hip/acc_utils.cpp
     acc/cuda_hip/acc_init.cpp
     acc/cuda_hip/acc_mem.cpp
     acc/cuda_hip/acc_stream.cpp)

diff --git a/src/acc/cuda_hip/acc_utils.cpp b/src/acc/cuda_hip/acc_utils.cpp
@@ -0,0 +1,25 @@
+/*------------------------------------------------------------------------------------------------*/
+/* Copyright (C) by the DBCSR developers group - All rights reserved                              */
+/* This file is part of the DBCSR library.                                                        */
+/*                                                                                                */
+/* For information on the license, see the LICENSE file.                                          */
+/* For further information please visit https://dbcsr.cp2k.org                                    */
+/* SPDX-License-Identifier: GPL-2.0+                                                              */
+/*------------------------------------------------------------------------------------------------*/
+
+#include "acc_utils.h"
+
+#if defined(__CUDA)
+#  include "../cuda/acc_cuda.h"
+#elif defined(__HIP)
+#  include "../hip/acc_hip.h"
+#endif
+
+//===========================================================================
+int acc_get_gpu_warp_size() {
+  int device = 0;
+  ACC(DeviceProp) prop;
+  ACC_API_CALL(GetDevice, (&device));
+  ACC_API_CALL(GetDeviceProperties, (&prop, device));
+  return prop.warpSize;
+}
diff --git a/src/acc/cuda_hip/acc_utils.h b/src/acc/cuda_hip/acc_utils.h
@@ -0,0 +1,15 @@
+/*------------------------------------------------------------------------------------------------*/
+/* Copyright (C) by the DBCSR developers group - All rights reserved                              */
+/* This file is part of the DBCSR library.                                                        */
+/*                                                                                                */
+/* For information on the license, see the LICENSE file.                                          */
+/* For further information please visit https://dbcsr.cp2k.org                                    */
+/* SPDX-License-Identifier: GPL-2.0+                                                              */
+/*------------------------------------------------------------------------------------------------*/
+
+#ifndef ACC_UTILS_H
+#define ACC_UTILS_H
+
+int acc_get_gpu_warp_size(void);
+
+#endif /*ACC_UTILS_H*/
diff --git a/src/acc/cuda_hip/calculate_norms.cpp b/src/acc/cuda_hip/calculate_norms.cpp
@@ -42,7 +42,7 @@
 #elif defined(__HIP)
 #  include "../hip/acc_hip.h"
 #endif
-#include "libsmm_acc_init.h"
+#include "acc_utils.h"
 
 template<int warpsz, int blocksz>
 __global__ void calculate_norms_d(

diff --git a/src/acc/libsmm_acc/libsmm_acc_init.cpp b/src/acc/libsmm_acc/libsmm_acc_init.cpp
@@ -9,6 +9,7 @@
 
 #include "libsmm_acc_init.h"
 #include "../acc_libsmm.h"
+#include "../cuda_hip/acc_utils.h"
 
 #if defined(_OPENMP)
 #  include <omp.h>
@@ -102,15 +103,6 @@ int libsmm_acc_check_gpu_warp_size_consistency() {
   return 0;
 }
 
-//===========================================================================
-int acc_get_gpu_warp_size() {
-  int device = 0;
-  ACC(DeviceProp) prop;
-  ACC_API_CALL(GetDevice, (&device));
-  ACC_API_CALL(GetDeviceProperties, (&prop, device));
-  return prop.warpSize;
-}
-
 //===========================================================================
 extern "C" int libsmm_acc_is_thread_safe() {
 #if defined(_OPENMP)

diff --git a/src/acc/libsmm_acc/libsmm_acc_init.h b/src/acc/libsmm_acc/libsmm_acc_init.h
@@ -25,8 +25,6 @@ int libsmm_acc_gpu_blas_init();
 
 int libsmm_acc_check_gpu_warp_size_consistency(void);
 
-int acc_get_gpu_warp_size(void);
-
 extern "C" int libsmm_acc_is_thread_safe(void);
 
 extern std::vector<ACC_BLAS(Handle_t)*> acc_blashandles;