cuda_shared_qualifier.h

/******************************************************************************
 * Copyright 2023 NVIDIA Corporation. All rights reserved.
 *****************************************************************************/
 
#ifndef NVIDIA_INDEX_CUDA_SHARED_QUALIFIER_H
#define NVIDIA_INDEX_CUDA_SHARED_QUALIFIER_H
 
// force inline
#if defined(__CUDACC__) || defined(__CUDABE__)
#define NV_IDX_FORCEINLINE __forceinline__
#elif defined(__GNUC__)
#define NV_IDX_FORCEINLINE inline 
#elif defined(_WIN32)
#define NV_IDX_FORCEINLINE __forceinline
#else
#define NV_IDX_FORCEINLINE inline
#endif
 
// noinline
#if defined(__CUDACC__) || defined(__CUDABE__)
#define NV_IDX_NOINLINE __noinline__
#elif defined(__GNUC__)
#define NV_IDX_NOINLINE static 
#elif defined(_WIN32) && defined(_MSC_VER)
#define NV_IDX_NOINLINE static __declspec(noinline)
#else
#define NV_IDX_NOINLINE static
#endif
 
#if defined(__CUDACC__) || defined(__CUDABE__)
 
// CUDA code path
#define NV_IDX_DEVICE_INLINE            __device__ __forceinline__    // REVIEW THIS, should not be forceinline
#define NV_IDX_DEVICE_NOINLINE          __device__ __noinline__
#define NV_IDX_DEVICE_FORCEINLINE       __device__ __forceinline__
 
#define NV_IDX_DEVICE_HOST_INLINE       __host__ NV_IDX_DEVICE_INLINE
#define NV_IDX_DEVICE_HOST_NOINLINE     __host__ NV_IDX_DEVICE_NOINLINE
#define NV_IDX_DEVICE_HOST_FORCEINLINE  __host__ NV_IDX_DEVICE_FORCEINLINE
 
#define NV_IDX_HOST_INLINE              __host__ inline
#define NV_IDX_HOST_NOINLINE            __host__ NV_IDX_NOINLINE
#define NV_IDX_HOST_FORCEINLINE         __host__ NV_IDX_FORCEINLINE
 
#define NV_IDX_CONSTANT                 static __device__ __constant__ 
 
#else
 
// CPU code path
#define NV_IDX_DEVICE_INLINE            inline
#define NV_IDX_DEVICE_NOINLINE          NV_IDX_NOINLINE
#define NV_IDX_DEVICE_FORCEINLINE       NV_IDX_FORCEINLINE
 
#define NV_IDX_DEVICE_HOST_INLINE       inline
#define NV_IDX_DEVICE_HOST_NOINLINE     NV_IDX_NOINLINE
#define NV_IDX_DEVICE_HOST_FORCEINLINE  NV_IDX_FORCEINLINE
 
#define NV_IDX_HOST_INLINE              inline
#define NV_IDX_HOST_NOINLINE            NV_IDX_NOINLINE
#define NV_IDX_HOST_FORCEINLINE         NV_IDX_FORCEINLINE
 
#define NV_IDX_CONSTANT                 static const
 
#endif // __CUDACC__
 
// derived definitions:
#define NV_IDX_DEVICE_INLINE_MEMBER     NV_IDX_DEVICE_INLINE
 
// CUDA requires 16 byte alignment. This shall be used for all structs that are shared between host
// and device.
#define NV_IDX_DEVICE_ALIGN alignas(16)
 
#endif // NVIDIA_INDEX_CUDA_SHARED_QUALIFIER_H