Desul atomics: cleanup macro guards in CUDA/HIP lock guard files

sandialabs · Feb 1, 2023 · 37bcd41 · 37bcd41
1 parent 23e2d85
commit 37bcd41
Show file tree

Hide file tree

Showing 4 changed files with 4 additions and 38 deletions.
diff --git a/tpls/desul/include/desul/atomics/Lock_Array_CUDA.hpp b/tpls/desul/include/desul/atomics/Lock_Array_CUDA.hpp
@@ -9,13 +9,11 @@ SPDX-License-Identifier: (BSD-3-Clause)
 #ifndef DESUL_ATOMICS_LOCK_ARRAY_CUDA_HPP_
 #define DESUL_ATOMICS_LOCK_ARRAY_CUDA_HPP_
 
+#include <cstdint>
+
 #include "desul/atomics/Common.hpp"
 #include "desul/atomics/Macros.hpp"
 
-#ifdef DESUL_HAVE_CUDA_ATOMICS
-
-#include <cstdint>
-
 namespace desul {
 namespace Impl {
 
@@ -42,14 +40,6 @@ void init_lock_arrays_cuda();
 template <typename /*AlwaysInt*/ = int>
 void finalize_lock_arrays_cuda();
 
-}  // namespace Impl
-}  // namespace desul
-
-#if defined(__CUDACC__)
-
-namespace desul {
-namespace Impl {
-
 /// \brief This global variable in CUDA space is what kernels use
 ///        to get access to the lock arrays.
 ///
@@ -118,12 +108,7 @@ __device__ inline void unlock_address_cuda(void* ptr, desul::MemoryScopeNode) {
   atomicExch(&desul::Impl::CUDA_SPACE_ATOMIC_LOCKS_NODE[offset], 0);
 }
 
-}  // namespace Impl
-}  // namespace desul
-
 // Make lock_array_copied an explicit translation unit scope thingy
-namespace desul {
-namespace Impl {
 namespace {
 static int lock_array_copied = 0;
 }  // namespace
@@ -149,13 +134,9 @@ inline static
 }  // namespace Impl
 }  // namespace desul
 
-#endif /* defined( __CUDACC__ ) */
-
-#endif /* defined( DESUL_HAVE_CUDA_ATOMICS ) */
-
 namespace desul {
 
-#if defined(__CUDACC_RDC__) || (!defined(__CUDACC__))
+#if defined(__CUDACC_RDC__)
 inline void ensure_cuda_lock_arrays_on_device() {}
 #else
 static inline void ensure_cuda_lock_arrays_on_device() {

diff --git a/tpls/desul/include/desul/atomics/Lock_Array_HIP.hpp b/tpls/desul/include/desul/atomics/Lock_Array_HIP.hpp
@@ -43,12 +43,6 @@ void init_lock_arrays_hip();
 ///   snapshotted version while also linking against pure Desul
 template <typename /*AlwaysInt*/ = int>
 void finalize_lock_arrays_hip();
-}  // namespace Impl
-}  // namespace desul
-
-#ifdef __HIPCC__
-namespace desul {
-namespace Impl {
 
 /**
  * \brief This global variable in HIP space is what kernels use to get access
@@ -120,13 +114,8 @@ __device__ inline void unlock_address_hip(void* ptr, desul::MemoryScopeNode) {
   offset = offset & HIP_SPACE_ATOMIC_MASK;
   atomicExch(&desul::Impl::HIP_SPACE_ATOMIC_LOCKS_NODE[offset], 0);
 }
-#endif
-}  // namespace Impl
-}  // namespace desul
 
 // Make lock_array_copied an explicit translation unit scope thing
-namespace desul {
-namespace Impl {
 namespace {
 static int lock_array_copied = 0;
 }  // namespace
@@ -150,7 +139,7 @@ inline static
 }
 }  // namespace Impl
 
-#if defined(DESUL_HIP_RDC) || (!defined(__HIPCC__))
+#if defined(DESUL_HIP_RDC)
 inline void ensure_hip_lock_arrays_on_device() {}
 #else
 static inline void ensure_hip_lock_arrays_on_device() {

diff --git a/tpls/desul/src/Lock_Array_CUDA.cpp b/tpls/desul/src/Lock_Array_CUDA.cpp
@@ -11,7 +11,6 @@ SPDX-License-Identifier: (BSD-3-Clause)
 #include <sstream>
 #include <string>
 
-#ifdef DESUL_HAVE_CUDA_ATOMICS
 #ifdef __CUDACC_RDC__
 namespace desul {
 namespace Impl {
@@ -96,4 +95,3 @@ template void finalize_lock_arrays_cuda<int>();
 }  // namespace Impl
 
 }  // namespace desul
-#endif
diff --git a/tpls/desul/src/Lock_Array_HIP.cpp b/tpls/desul/src/Lock_Array_HIP.cpp
@@ -11,7 +11,6 @@ SPDX-License-Identifier: (BSD-3-Clause)
 #include <sstream>
 #include <string>
 
-#ifdef DESUL_HAVE_HIP_ATOMICS
 #ifdef DESUL_HIP_RDC
 namespace desul {
 namespace Impl {
@@ -99,4 +98,3 @@ template void finalize_lock_arrays_hip<int>();
 }  // namespace Impl
 
 }  // namespace desul
-#endif