include/VecCore/CUDA.h

0001 #ifndef VECCORE_CUDA_H
0002 #define VECCORE_CUDA_H
0003
0004 #if !defined(VECCORE_ENABLE_CUDA)
0005
0006 #define VECCORE_IMPL_NAMESPACE cxx
0007
0008 #define VECCORE_ATT_HOST               /* empty */
0009 #define VECCORE_ATT_DEVICE             /* empty */
0010 #define VECCORE_ATT_HOST_DEVICE        /* empty */
0011 #define VECCORE_ATT_GLOBAL             /* empty */
0012 #define VECCORE_CUDA_ALIGN             /* empty */
0013 #define VECCORE_DECLARE_CUDA(x)        /* empty */
0014 #define VECCORE_DECLARE_CUDA_TYPE(T)   /* empty */
0015 #define VECCORE_DECLARE_CUDA_CLASS(x)  /* empty */
0016 #define VECCORE_DECLARE_CUDA_STRUCT(x) /* empty */
0017 #define VECCORE_DEVICE_CONSTANT
0018
0019 #elif (defined(__CUDACC__) || defined(__NVCC__))
0020
0021 #define VECCORE_IMPL_NAMESPACE cuda
0022
0023 #define VECCORE_CUDA
0024 #ifdef __CUDA_ARCH__
0025 #define VECCORE_CUDA_DEVICE_COMPILATION
0026 #endif
0027
0028 #define VECCORE_ATT_HOST __host__
0029 #define VECCORE_ATT_DEVICE __device__
0030 #define VECCORE_ATT_HOST_DEVICE __host__ __device__
0031 #define VECCORE_ATT_GLOBAL __global__
0032 #define VECCORE_CUDA_ALIGN __align__((64))
0033 #define VECCORE_DECLARE_CUDA(x)          /* empty */
0034 #define VECCORE_DECLARE_CUDA_TYPE(T)     /* empty */
0035 #define VECCORE_DECLARE_CUDA_CLASS(x)    /* empty */
0036 #define VECCORE_DECLARE_CUDA_STRUCT(x)   /* empty */
0037 #define VECCORE_DECLARE_CUDA_TEMPLATE(x) /* empty */
0038 #define VECCORE_DEVICE_CONSTANT __constant__
0039
0040 #else // CUDA enabled, but compiling regular C++ code
0041
0042 #define VECCORE_CUDA_INTERFACE
0043 #define VECCORE_ATT_HOST        /* empty */
0044 #define VECCORE_ATT_DEVICE      /* empty */
0045 #define VECCORE_ATT_HOST_DEVICE /* empty */
0046 #define VECCORE_ATT_GLOBAL      /* empty */
0047 #define VECCORE_CUDA_ALIGN      /* empty */
0048
0049 // Keep the macro compact
0050 // clang-format off
0051
0052 #define VECCORE_CUDA_BASIC_TYPES                                               \
0053 template <typename T> struct CudaTypeTraits;                                   \
0054 template <> struct CudaTypeTraits<float>    { using Type = float;    };        \
0055 template <> struct CudaTypeTraits<double>   { using Type = double;   };        \
0056 template <> struct CudaTypeTraits<int8_t>   { using Type = int8_t;   };        \
0057 template <> struct CudaTypeTraits<int16_t>  { using Type = int16_t;  };        \
0058 template <> struct CudaTypeTraits<int32_t>  { using Type = int32_t;  };        \
0059 template <> struct CudaTypeTraits<int64_t>  { using Type = int64_t;  };        \
0060 template <> struct CudaTypeTraits<uint8_t>  { using Type = uint8_t;  };        \
0061 template <> struct CudaTypeTraits<uint16_t> { using Type = uint16_t; };        \
0062 template <> struct CudaTypeTraits<uint32_t> { using Type = uint32_t; };        \
0063 template <> struct CudaTypeTraits<uint64_t> { using Type = uint64_t; };        \
0064 template <typename T> using CudaType = typename CudaTypeTraits<T>::Type;
0065
0066 #define VECCORE_DECLARE_CUDA(T) T; namespace cuda { T; }
0067
0068 #define VECCORE_DECLARE_CUDA_TYPE(T)                                           \
0069   template <> struct CudaTypeTraits<T> { using Type = cuda::T; };              \
0070
0071 #define VECCORE_DECLARE_CUDA_CLASS(x)                                          \
0072   VECCORE_DECLARE_CUDA(class x) VECCORE_DECLARE_CUDA_TYPE(x)
0073
0074 #define VECCORE_DECLARE_CUDA_STRUCT(x)                                         \
0075   VECCORE_DECLARE_CUDA(struct x) VECCORE_DECLARE_CUDA_TYPE(x)
0076
0077 // clang-format on
0078
0079 #endif // defined (VECCORE_ENABLE_CUDA)
0080
0081 #endif