Utilities/interface/DeviceGlobal.h

0001 #ifndef FWCore_Utilities_DeviceGlobal_h
0002 #define FWCore_Utilities_DeviceGlobal_h
0003
0004 // FIXME alpaka provides ALPAKA_STATIC_ACC_MEM_GLOBAL to declare device global
0005 // variables, but it is currently not working as expected. Improve its behaviour
0006 // and syntax and migrate to that.
0007
0008 #if defined(__SYCL_DEVICE_ONLY__)
0009
0010 // The SYCL standard does not support device global variables.
0011 // oneAPI defines the sycl_ext_oneapi_device_global extension, but with an awkward syntax
0012 // that is not easily compatible with CUDA, HIP and regular C++ global variables.
0013 #error "The SYCL backend does not support device global variables"
0014 #define DEVICE_GLOBAL
0015
0016 #elif defined(__CUDA_ARCH__) or defined(__HIP_DEVICE_COMPILE__)
0017
0018 // CUDA and HIP/ROCm device compilers use the __device__ attribute.
0019 #define DEVICE_GLOBAL __device__
0020
0021 #else
0022
0023 // host compilers do not need any special attributes.
0024 #define DEVICE_GLOBAL
0025
0026 #endif
0027
0028 #endif  // FWCore_Utilities_DeviceGlobal_h