Skip to content

Commit

Permalink
Include use of NVHPC in CUB/Thrust magic namespace (#2771)
Browse files Browse the repository at this point in the history 
The new Thrust namespace is `thrust::THRUST_200800_SM_860_NVHPC_NS` if NVHPC is used, otherwise `thrust::THRUST_200800_SM_860_NS`. CUB works analogous.

Fixes: #2770
  • Loading branch information
@bernhardmgruber
bernhardmgruber authored Nov 19, 2024
1 parent 052db85 commit f9d56e5
Show file tree
Hide file tree
Showing 3 changed files with 49 additions and 110 deletions.
59 changes: 6 additions & 53 deletions cub/cub/util_namespace.cuh
View file
Original file line number Diff line number Diff line change
Expand Up @@ -117,56 +117,6 @@
# define CUB_NS_QUALIFIER ::cub
#endif

#if !defined(CUB_DETAIL_MAGIC_NS_NAME)
# define CUB_DETAIL_COUNT_N( \
_1, _2, _3, _4, _5, _6, _7, _8, _9, _10, _11, _12, _13, _14, _15, _16, _17, _18, _19, _20, N, ...) \
N
# define CUB_DETAIL_COUNT(...) \
CUB_DETAIL_IDENTITY( \
CUB_DETAIL_COUNT_N(__VA_ARGS__, 20, 19, 18, 17, 16, 15, 14, 13, 12, 11, 10, 9, 8, 7, 6, 5, 4, 3, 2, 1))
# define CUB_DETAIL_IDENTITY(N) N
# define CUB_DETAIL_APPLY(MACRO, ...) CUB_DETAIL_IDENTITY(MACRO(__VA_ARGS__))
# define CUB_DETAIL_MAGIC_NS_NAME1(P1) CUB_##P1##_NS
# define CUB_DETAIL_MAGIC_NS_NAME2(P1, P2) CUB_##P1##_##P2##_NS
# define CUB_DETAIL_MAGIC_NS_NAME3(P1, P2, P3) CUB_##P1##_##P2##_##P3##_NS
# define CUB_DETAIL_MAGIC_NS_NAME4(P1, P2, P3, P4) CUB_##P1##_##P2##_##P3##_##P4##_NS
# define CUB_DETAIL_MAGIC_NS_NAME5(P1, P2, P3, P4, P5) CUB_##P1##_##P2##_##P3##_##P4##_##P5##_NS
# define CUB_DETAIL_MAGIC_NS_NAME6(P1, P2, P3, P4, P5, P6) CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_NS
# define CUB_DETAIL_MAGIC_NS_NAME7(P1, P2, P3, P4, P5, P6, P7) CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_NS
# define CUB_DETAIL_MAGIC_NS_NAME8(P1, P2, P3, P4, P5, P6, P7, P8) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_NS
# define CUB_DETAIL_MAGIC_NS_NAME9(P1, P2, P3, P4, P5, P6, P7, P8, P9) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_NS
# define CUB_DETAIL_MAGIC_NS_NAME10(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_NS
# define CUB_DETAIL_MAGIC_NS_NAME11(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_NS
# define CUB_DETAIL_MAGIC_NS_NAME12(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_NS
# define CUB_DETAIL_MAGIC_NS_NAME13(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_NS
# define CUB_DETAIL_MAGIC_NS_NAME14(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_NS
# define CUB_DETAIL_MAGIC_NS_NAME15(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_NS
# define CUB_DETAIL_MAGIC_NS_NAME16(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_NS
# define CUB_DETAIL_MAGIC_NS_NAME17(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_NS
# define CUB_DETAIL_MAGIC_NS_NAME18(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_NS
# define CUB_DETAIL_MAGIC_NS_NAME19( \
P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18, P19) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_##P19##_NS
# define CUB_DETAIL_MAGIC_NS_NAME20( \
P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18, P19, P20) \
CUB_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_##P19##_##P20##_NS
# define CUB_DETAIL_DISPATCH(N) CUB_DETAIL_MAGIC_NS_NAME##N
# define CUB_DETAIL_MAGIC_NS_NAME(...) \
CUB_DETAIL_IDENTITY(CUB_DETAIL_APPLY(CUB_DETAIL_DISPATCH, CUB_DETAIL_COUNT(__VA_ARGS__))(__VA_ARGS__))
#endif // !defined(CUB_DETAIL_MAGIC_NS_NAME)

// clang-format off
#if defined(CUB_DISABLE_NAMESPACE_MAGIC) || defined(CUB_WRAPPED_NAMESPACE)
# if !defined(CUB_WRAPPED_NAMESPACE)
# if !defined(CUB_IGNORE_NAMESPACE_MAGIC_ERROR)
Expand All @@ -177,14 +127,17 @@
# define CUB_DETAIL_MAGIC_NS_END
#else // not defined(CUB_DISABLE_NAMESPACE_MAGIC)
# if defined(_NVHPC_CUDA)
# define CUB_DETAIL_MAGIC_NS_BEGIN inline namespace CUB_DETAIL_MAGIC_NS_NAME(CUB_VERSION, NV_TARGET_SM_INTEGER_LIST) {
# define CUB_DETAIL_MAGIC_NS_BEGIN \
inline namespace _CCCL_PP_SPLICE_WITH(_, CUB, CUB_VERSION, SM, NV_TARGET_SM_INTEGER_LIST, NVHPC) \
{
# define CUB_DETAIL_MAGIC_NS_END }
# else // not defined(_NVHPC_CUDA)
# define CUB_DETAIL_MAGIC_NS_BEGIN inline namespace CUB_DETAIL_MAGIC_NS_NAME(CUB_VERSION, __CUDA_ARCH_LIST__) {
# define CUB_DETAIL_MAGIC_NS_BEGIN \
inline namespace _CCCL_PP_SPLICE_WITH(_, CUB, CUB_VERSION, SM, __CUDA_ARCH_LIST__) \
{
# define CUB_DETAIL_MAGIC_NS_END }
# endif // not defined(_NVHPC_CUDA)
#endif // not defined(CUB_DISABLE_NAMESPACE_MAGIC)
// clang-format on

/**
* \def CUB_NAMESPACE_BEGIN
Expand Down
37 changes: 37 additions & 0 deletions libcudacxx/include/cuda/std/__cccl/preprocessor.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -1127,4 +1127,41 @@
#define _CCCL_PP_REPEAT_REVERSE255(_M, _S, _F) _CCCL_PP_REPEAT_REVERSE254(_M, _F(_S), _F) _M(_S)
#define _CCCL_PP_REPEAT_REVERSE256(_M, _S, _F) _CCCL_PP_REPEAT_REVERSE255(_M, _F(_S), _F) _M(_S)

#define _CCCL_PP_SPLICE_WITH_IMPL1(SEP, P1) P1
#define _CCCL_PP_SPLICE_WITH_IMPL2(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL1(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL3(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL2(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL4(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL3(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL5(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL4(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL6(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL5(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL7(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL6(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL8(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL7(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL9(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL8(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL10(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL9(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL11(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL10(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL12(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL11(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL13(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL12(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL14(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL13(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL15(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL14(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL16(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL15(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL17(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL16(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL18(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL17(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL19(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL18(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL21(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL19(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL22(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL21(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL23(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL22(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL24(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL23(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL25(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL24(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL26(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL25(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL27(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL26(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL28(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL27(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL29(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL28(SEP, __VA_ARGS__))
#define _CCCL_PP_SPLICE_WITH_IMPL30(SEP, P1, ...) _CCCL_PP_CAT(P1##SEP, _CCCL_PP_SPLICE_WITH_IMPL29(SEP, __VA_ARGS__))

#define _CCCL_PP_SPLICE_WITH_IMPL_DISPATCH(N) _CCCL_PP_SPLICE_WITH_IMPL##N

// Splices a pack of arguments into a single token, separated by SEP
// E.g., _CCCL_PP_SPLICE_WITH(_, A, B, C) will evaluate to A_B_C
#define _CCCL_PP_SPLICE_WITH(SEP, ...) \
_CCCL_PP_EXPAND(_CCCL_PP_EVAL(_CCCL_PP_SPLICE_WITH_IMPL_DISPATCH, _CCCL_PP_COUNT(__VA_ARGS__))(SEP, __VA_ARGS__))

#endif // __CCCL_PREPROCESSOR_H
63 changes: 6 additions & 57 deletions thrust/thrust/detail/config/namespace.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -97,61 +97,7 @@
# define THRUST_NS_QUALIFIER ::thrust
#endif

// clang-format off
#if THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
# if !defined(THRUST_DETAIL_ABI_NS_NAME)
# define THRUST_DETAIL_COUNT_N(_1, _2, _3, _4, _5, _6, _7, _8, _9, _10, _11, _12, _13, \
_14, _15, _16, _17, _18, _19, _20, N, ...) \
N
# define THRUST_DETAIL_COUNT(...) \
THRUST_DETAIL_IDENTITY(THRUST_DETAIL_COUNT_N(__VA_ARGS__, 20, 19, 18, 17, 16, 15, 14, 13, 12, \
11, 10, 9, 8, 7, 6, 5, 4, 3, 2, 1))
# define THRUST_DETAIL_IDENTITY(N) N
# define THRUST_DETAIL_APPLY(MACRO, ...) THRUST_DETAIL_IDENTITY(MACRO(__VA_ARGS__))
# define THRUST_DETAIL_ABI_NS_NAME1(P1) \
THRUST_##P1##_NS
# define THRUST_DETAIL_ABI_NS_NAME2(P1, P2) \
THRUST_##P1##_##P2##_NS
# define THRUST_DETAIL_ABI_NS_NAME3(P1, P2, P3) \
THRUST_##P1##_##P2##_##P3##_NS
# define THRUST_DETAIL_ABI_NS_NAME4(P1, P2, P3, P4) \
THRUST_##P1##_##P2##_##P3##_##P4##_NS
# define THRUST_DETAIL_ABI_NS_NAME5(P1, P2, P3, P4, P5) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_NS
# define THRUST_DETAIL_ABI_NS_NAME6(P1, P2, P3, P4, P5, P6) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_NS
# define THRUST_DETAIL_ABI_NS_NAME7(P1, P2, P3, P4, P5, P6, P7) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_NS
# define THRUST_DETAIL_ABI_NS_NAME8(P1, P2, P3, P4, P5, P6, P7, P8) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_NS
# define THRUST_DETAIL_ABI_NS_NAME9(P1, P2, P3, P4, P5, P6, P7, P8, P9) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_NS
# define THRUST_DETAIL_ABI_NS_NAME10(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_NS
# define THRUST_DETAIL_ABI_NS_NAME11(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_NS
# define THRUST_DETAIL_ABI_NS_NAME12(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_NS
# define THRUST_DETAIL_ABI_NS_NAME13(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_NS
# define THRUST_DETAIL_ABI_NS_NAME14(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_NS
# define THRUST_DETAIL_ABI_NS_NAME15(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_NS
# define THRUST_DETAIL_ABI_NS_NAME16(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_NS
# define THRUST_DETAIL_ABI_NS_NAME17(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_NS
# define THRUST_DETAIL_ABI_NS_NAME18(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_NS
# define THRUST_DETAIL_ABI_NS_NAME19(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18, P19) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_##P19##_NS
# define THRUST_DETAIL_ABI_NS_NAME20(P1, P2, P3, P4, P5, P6, P7, P8, P9, P10, P11, P12, P13, P14, P15, P16, P17, P18, P19, P20) \
THRUST_##P1##_##P2##_##P3##_##P4##_##P5##_##P6##_##P7##_##P8##_##P9##_##P10##_##P11##_##P12##_##P13##_##P14##_##P15##_##P16##_##P17##_##P18##_##P19##_##P20##_NS
# define THRUST_DETAIL_DISPATCH(N) THRUST_DETAIL_ABI_NS_NAME ## N
# define THRUST_DETAIL_ABI_NS_NAME(...) THRUST_DETAIL_IDENTITY(THRUST_DETAIL_APPLY(THRUST_DETAIL_DISPATCH, THRUST_DETAIL_COUNT(__VA_ARGS__))(__VA_ARGS__))
# endif // !defined(THRUST_DETAIL_ABI_NS_NAME)

# if defined(THRUST_DISABLE_ABI_NAMESPACE) || defined(THRUST_WRAPPED_NAMESPACE)
# if !defined(THRUST_WRAPPED_NAMESPACE)
# if !defined(THRUST_IGNORE_ABI_NAMESPACE_ERROR)
Expand All @@ -162,18 +108,21 @@
# define THRUST_DETAIL_ABI_NS_END
# else // not defined(THRUST_DISABLE_ABI_NAMESPACE)
# if defined(_NVHPC_CUDA)
# define THRUST_DETAIL_ABI_NS_BEGIN inline namespace THRUST_DETAIL_ABI_NS_NAME(THRUST_VERSION, NV_TARGET_SM_INTEGER_LIST) {
# define THRUST_DETAIL_ABI_NS_BEGIN \
inline namespace _CCCL_PP_SPLICE_WITH(_, THRUST, THRUST_VERSION, SM, NV_TARGET_SM_INTEGER_LIST, NVHPC, NS) \
{
# define THRUST_DETAIL_ABI_NS_END }
# else // not defined(_NVHPC_CUDA)
# define THRUST_DETAIL_ABI_NS_BEGIN inline namespace THRUST_DETAIL_ABI_NS_NAME(THRUST_VERSION, __CUDA_ARCH_LIST__) {
# define THRUST_DETAIL_ABI_NS_BEGIN \
inline namespace _CCCL_PP_SPLICE_WITH(_, THRUST, THRUST_VERSION, SM, __CUDA_ARCH_LIST__, NS) \
{
# define THRUST_DETAIL_ABI_NS_END }
# endif // not defined(_NVHPC_CUDA)
# endif // not defined(THRUST_DISABLE_ABI_NAMESPACE)
#else // THRUST_DEVICE_SYSTEM != THRUST_DEVICE_SYSTEM_CUDA
# define THRUST_DETAIL_ABI_NS_BEGIN
# define THRUST_DETAIL_ABI_NS_END
#endif // THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
// clang-format on

/**
* \def THRUST_NAMESPACE_BEGIN
Expand Down

0 comments on commit f9d56e5

Please sign in to comment.