Skip to content

Backport - Make compiler version comparisons safer (#4185) #4394

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
miscco merged 2 commits into from
Apr 10, 2025
Merged

Backport - Make compiler version comparisons safer (#4185) #4394

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
c518f22
Make compiler version comparisons safer (#4185)
davebayer Mar 21, 2025
956528e
Fix compile issues
miscco Apr 10, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
4 changes: 2 additions & 2 deletions cub/cub/util_device.cuh
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -688,6 +688,6 @@ private:

CUB_NAMESPACE_END

#if _CCCL_HAS_CUDA_COMPILER && !_CCCL_COMPILER(NVRTC)
#if _CCCL_HAS_CUDA_COMPILER() && !_CCCL_COMPILER(NVRTC)
# include <cub/detail/launcher/cuda_runtime.cuh> // to complete the definition of TripleChevronFactory
#endif // _CCCL_HAS_CUDA_COMPILER && !_CCCL_COMPILER(NVRTC)
#endif // _CCCL_HAS_CUDA_COMPILER() && !_CCCL_COMPILER(NVRTC)
2 changes: 1 addition & 1 deletion cudax/include/cuda/experimental/__detail/config.cuh
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -24,7 +24,7 @@
// Debuggers do not step into functions marked with __attribute__((__artificial__)).
// This is useful for small wrapper functions that just dispatch to other functions and
// that are inlined into the caller.
#if _CCCL_HAS_ATTRIBUTE(__artificial__) && !_CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_ATTRIBUTE(__artificial__) && !_CCCL_HAS_CUDA_COMPILER()
# define _CUDAX_ARTIFICIAL __attribute__((__artificial__))
#else // ^^^ _CCCL_HAS_ATTRIBUTE(__artificial__) ^^^ / vvv !_CCCL_HAS_ATTRIBUTE(__artificial__) vvv
# define _CUDAX_ARTIFICIAL
Expand Down
2 changes: 1 addition & 1 deletion docs/cccl/development/macro.rst
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -59,7 +59,7 @@ The ``_CCCL_CUDA_COMPILER`` function-like macro can also be used to check the ve
**CUDA identification/version macros**:

+----------------------------------+-----------------------------+
| ``_CCCL_HAS_CUDA_COMPILER`` | CUDA compiler is available |
| ``_CCCL_HAS_CUDA_COMPILER()`` | CUDA compiler is available |
+----------------------------------+-----------------------------+
| ``_CCCL_CUDACC_BELOW(12, 7)`` | CUDA version below 12.7 |
+----------------------------------+-----------------------------+
Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/codegen/generators/header.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -57,7 +57,7 @@ inline void FormatHeader(std::ostream& out)

_LIBCUDACXX_BEGIN_NAMESPACE_STD

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
)XXX";

out << header;
Expand All @@ -66,7 +66,7 @@ _LIBCUDACXX_BEGIN_NAMESPACE_STD
inline void FormatTail(std::ostream& out)
{
std::string tail = R"XXX(
#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_END_NAMESPACE_STD

Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__barrier/barrier_arrive_tx.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# if __cccl_ptx_isa >= 800

# include <cuda/__barrier/barrier_block_scope.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__barrier/barrier_block_scope.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@

#include <cuda/__fwd/barrier.h>
#include <cuda/__fwd/barrier_native_handle.h>
#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# include <cuda/__ptx/instructions/mbarrier_arrive.h>
# include <cuda/__ptx/ptx_dot_variants.h>
# include <cuda/__ptx/ptx_helper_functions.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__barrier/barrier_expect_tx.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# if __cccl_ptx_isa >= 800

# include <cuda/__barrier/barrier_block_scope.h>
Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/__barrier/barrier_native_handle.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -25,7 +25,7 @@
#include <cuda/__fwd/barrier.h>
#include <cuda/std/cstdint>

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_BEGIN_NAMESPACE_CUDA_DEVICE

Expand All @@ -36,6 +36,6 @@ _CCCL_DEVICE inline _CUDA_VSTD::uint64_t* barrier_native_handle(barrier<thread_s

_LIBCUDACXX_END_NAMESPACE_CUDA_DEVICE

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

#endif // _CUDA___BARRIER_BARRIER_NATIVE_HANDLE_H
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/__bit/bit_reverse.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -60,7 +60,7 @@ _CCCL_NODISCARD _LIBCUDACXX_HIDE_FROM_ABI constexpr _Tp __bit_reverse_builtin(_T

#endif // defined(_CCCL_BUILTIN_BITREVERSE32)

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

template <typename _Tp>
_CCCL_NODISCARD _CCCL_HIDE_FROM_ABI _CCCL_DEVICE constexpr _Tp __bit_reverse_device(_Tp __value) noexcept
Expand Down Expand Up @@ -92,7 +92,7 @@ _CCCL_NODISCARD _CCCL_HIDE_FROM_ABI _CCCL_DEVICE constexpr _Tp __bit_reverse_dev
_CCCL_UNREACHABLE();
}

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

template <typename _Tp>
_CCCL_NODISCARD _LIBCUDACXX_HIDE_FROM_ABI constexpr _Tp __bit_reverse_generic(_Tp __value) noexcept
Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/__functional/for_each_canceled.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -28,7 +28,7 @@

#include <nv/target>

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_BEGIN_NAMESPACE_CUDA

Expand Down Expand Up @@ -270,6 +270,6 @@ _CCCL_DEVICE _CCCL_HIDE_FROM_ABI void for_each_canceled_block(__UnaryFunction __

_LIBCUDACXX_END_NAMESPACE_CUDA

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

#endif // _CUDA__FUNCTIONAL_FOR_EACH_CANCELED_H
6 changes: 3 additions & 3 deletions libcudacxx/include/cuda/__functional/get_device_address.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -38,7 +38,7 @@ _LIBCUDACXX_BEGIN_NAMESPACE_CUDA
template <class _Tp>
_CCCL_NODISCARD _LIBCUDACXX_HIDE_FROM_ABI _Tp* get_device_address(_Tp& __device_object)
{
#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
NV_IF_ELSE_TARGET(
NV_IS_DEVICE,
(return _CUDA_VSTD::addressof(__device_object);),
Expand All @@ -48,9 +48,9 @@ _CCCL_NODISCARD _LIBCUDACXX_HIDE_FROM_ABI _Tp* get_device_address(_Tp& __device_
&__device_ptr,
__device_object);
return static_cast<_Tp*>(__device_ptr);))
#else // ^^^ _CCCL_HAS_CUDA_COMPILER ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER vvv
#else // ^^^ _CCCL_HAS_CUDA_COMPILER() ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER() vvv
return _CUDA_VSTD::addressof(__device_object);
#endif // !_CCCL_HAS_CUDA_COMPILER
#endif // !_CCCL_HAS_CUDA_COMPILER()
}

_LIBCUDACXX_END_NAMESPACE_CUDA
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__memcpy_async/cp_async_bulk_shared_global.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# if __cccl_ptx_isa >= 800

# include <cuda/__ptx/instructions/cp_async_bulk.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__memcpy_async/cp_async_shared_global.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

# include <cuda/__ptx/ptx_dot_variants.h>
# include <cuda/__ptx/ptx_helper_functions.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__memcpy_async/memcpy_async.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

# include <cuda/__barrier/aligned_size.h>
# include <cuda/__barrier/async_contract_fulfillment.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__memcpy_async/memcpy_async_tx.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,7 +22,7 @@
# pragma system_header
#endif // no system header

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# if __cccl_ptx_isa >= 800

# include <cuda/__barrier/aligned_size.h>
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/__memcpy_async/memcpy_completion.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -31,7 +31,7 @@
#include <cuda/std/__atomic/scopes.h>
#include <cuda/std/cstdint>

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# include <cuda/__ptx/ptx_dot_variants.h>
# include <cuda/__ptx/ptx_helper_functions.h>
#endif // _CCCL_CUDA_COMPILER
Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/__ptx/ptx_helper_functions.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -27,7 +27,7 @@
#include <cuda/std/cstddef>
#include <cuda/std/cstdint>

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_BEGIN_NAMESPACE_CUDA_PTX

Expand Down Expand Up @@ -142,6 +142,6 @@ inline _CCCL_DEVICE _B8 __u32_as_b8(uint32_t __u32)

_LIBCUDACXX_END_NAMESPACE_CUDA_PTX

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

#endif // _CUDA_PTX_HELPER_FUNCTIONS_H_
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/std/__atomic/functions/cuda_local.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -33,7 +33,7 @@
// uses inline PTX to bypass __isLocal.
_LIBCUDACXX_BEGIN_NAMESPACE_STD

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

_CCCL_DEVICE inline bool __cuda_is_local(const volatile void* __ptr)
{
Expand Down Expand Up @@ -203,7 +203,7 @@ _CCCL_DEVICE bool __cuda_fetch_min_weak_if_local(volatile _Type* __ptr, _Type __
return __cuda_fetch_weak_if_local(__ptr, __val, __ret, __cuda_fetch_local_bop_min<_Type>);
}

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_END_NAMESPACE_STD

Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/std/__atomic/functions/cuda_ptx_derived.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -29,7 +29,7 @@

_LIBCUDACXX_BEGIN_NAMESPACE_STD

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

template <class _Operand>
using __cuda_atomic_enable_non_native_arithmetic =
Expand Down Expand Up @@ -390,7 +390,7 @@ _CCCL_DEVICE static inline void __atomic_signal_fence_cuda(int)
asm volatile("" ::: "memory");
}

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_END_NAMESPACE_STD

Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/std/__atomic/functions/cuda_ptx_generated.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -40,7 +40,7 @@

_LIBCUDACXX_BEGIN_NAMESPACE_STD

#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()

static inline _CCCL_DEVICE void __cuda_atomic_membar(__thread_scope_block_tag)
{ asm volatile("membar.cta;" ::: "memory"); }
Expand Down Expand Up @@ -3958,7 +3958,7 @@ static inline _CCCL_DEVICE _Type __atomic_fetch_sub_cuda(_Type volatile* __ptr,
return __atomic_fetch_add_cuda(__ptr, -__op, __memorder, _Sco{});
}

#endif // _CCCL_HAS_CUDA_COMPILER
#endif // _CCCL_HAS_CUDA_COMPILER()

_LIBCUDACXX_END_NAMESPACE_STD

Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/std/__atomic/wait/notify_wait.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -54,7 +54,7 @@ _LIBCUDACXX_HIDE_FROM_ABI void __atomic_notify_all(_Tp const volatile*, _Sco)
template <typename _Tp>
_LIBCUDACXX_HIDE_FROM_ABI bool __nonatomic_compare_equal(_Tp const& __lhs, _Tp const& __rhs)
{
#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
return __lhs == __rhs;
#else
return _CUDA_VSTD::memcmp(&__lhs, &__rhs, sizeof(_Tp)) == 0;
Expand Down
12 changes: 6 additions & 6 deletions libcudacxx/include/cuda/std/__cccl/assert.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -100,11 +100,11 @@ _CCCL_HOST_DEVICE
_CCCL_BUILTIN_EXPECT(static_cast<bool>(expression), 1) \
? (void) 0 : __assert_fail(message, __FILE__, __LINE__, __func__)
# endif // !_CCCL_COMPILER(MSVC)
#elif _CCCL_HAS_CUDA_COMPILER
#elif _CCCL_HAS_CUDA_COMPILER()
# define _CCCL_ASSERT_IMPL_DEVICE(expression, message) _CCCL_ASSERT_IMPL_HOST(expression, message)
#else // ^^^ _CCCL_HAS_CUDA_COMPILER ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER vvv
#else // ^^^ _CCCL_HAS_CUDA_COMPILER() ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER() vvv
# define _CCCL_ASSERT_IMPL_DEVICE(expression, message) ((void) 0)
#endif // !_CCCL_HAS_CUDA_COMPILER
#endif // !_CCCL_HAS_CUDA_COMPILER()

//! _CCCL_ASSERT_HOST is enabled conditionally depending on CCCL_ENABLE_HOST_ASSERTIONS
#ifdef CCCL_ENABLE_HOST_ASSERTIONS
Expand All @@ -130,17 +130,17 @@ _CCCL_HOST_DEVICE
# else
# define _CCCL_ASSERT(expression, message) ((void) 0)
# endif
#elif _CCCL_HAS_CUDA_COMPILER
#elif _CCCL_HAS_CUDA_COMPILER()
# ifdef __CUDA_ARCH__
# define _CCCL_VERIFY(expression, message) _CCCL_ASSERT_IMPL_DEVICE(expression, message)
# define _CCCL_ASSERT(expression, message) _CCCL_ASSERT_DEVICE(expression, message)
# else // ^^^ __CUDA_ARCH__ ^^^ / vvv !__CUDA_ARCH__ vvv
# define _CCCL_VERIFY(expression, message) _CCCL_ASSERT_IMPL_HOST(expression, message)
# define _CCCL_ASSERT(expression, message) _CCCL_ASSERT_HOST(expression, message)
# endif // !__CUDA_ARCH__
#else // ^^^ _CCCL_HAS_CUDA_COMPILER ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER vvv
#else // ^^^ _CCCL_HAS_CUDA_COMPILER() ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER() vvv
# define _CCCL_VERIFY(expression, message) _CCCL_ASSERT_IMPL_HOST(expression, message)
# define _CCCL_ASSERT(expression, message) _CCCL_ASSERT_HOST(expression, message)
#endif // !_CCCL_HAS_CUDA_COMPILER
#endif // !_CCCL_HAS_CUDA_COMPILER()

#endif // __CCCL_ASSERT_H
8 changes: 4 additions & 4 deletions libcudacxx/include/cuda/std/__cccl/builtin.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -239,19 +239,19 @@
# undef _CCCL_BUILTIN_BSWAP128
#endif // _CCCL_CUDA_COMPILER(NVCC)

#if _CCCL_CHECK_BUILTIN(builtin_bitreverse8) && !_CCCL_HAS_CUDA_COMPILER
#if _CCCL_CHECK_BUILTIN(builtin_bitreverse8) && !_CCCL_HAS_CUDA_COMPILER()
# define _CCCL_BUILTIN_BITREVERSE8(...) __builtin_bitreverse8(__VA_ARGS__)
#endif

#if _CCCL_CHECK_BUILTIN(builtin_bitreverse16) && !_CCCL_HAS_CUDA_COMPILER
#if _CCCL_CHECK_BUILTIN(builtin_bitreverse16) && !_CCCL_HAS_CUDA_COMPILER()
# define _CCCL_BUILTIN_BITREVERSE16(...) __builtin_bitreverse16(__VA_ARGS__)
#endif

#if _CCCL_CHECK_BUILTIN(builtin_bitreverse32) && !_CCCL_HAS_CUDA_COMPILER
#if _CCCL_CHECK_BUILTIN(builtin_bitreverse32) && !_CCCL_HAS_CUDA_COMPILER()
# define _CCCL_BUILTIN_BITREVERSE32(...) __builtin_bitreverse32(__VA_ARGS__)
#endif

#if _CCCL_CHECK_BUILTIN(builtin_bitreverse64) && !_CCCL_HAS_CUDA_COMPILER
#if _CCCL_CHECK_BUILTIN(builtin_bitreverse64) && !_CCCL_HAS_CUDA_COMPILER()
# define _CCCL_BUILTIN_BITREVERSE64(...) __builtin_bitreverse64(__VA_ARGS__)
#endif

Expand Down
6 changes: 3 additions & 3 deletions libcudacxx/include/cuda/std/__cccl/compiler.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -139,12 +139,12 @@
#define _CCCL_CUDACC_AT_LEAST(...) _CCCL_VERSION_COMPARE(_CCCL_CUDACC_, _CCCL_CUDACC, >=, __VA_ARGS__)

#if _CCCL_VERSION_IS_INVALID(_CCCL_CUDACC())
# define _CCCL_HAS_CUDA_COMPILER 0
# define _CCCL_HAS_CUDA_COMPILER() 0
#else // ^^^ has cuda compiler ^^^ / vvv no cuda compiler vvv
# define _CCCL_HAS_CUDA_COMPILER 1
# define _CCCL_HAS_CUDA_COMPILER() 1
#endif // ^^^ no cuda compiler ^^^

#if defined(_CCCL_HAS_CUDA_COMPILER) && _CCCL_CUDACC_BELOW(12) && !defined(CCCL_IGNORE_DEPRECATED_CUDA_BELOW_12)
#if _CCCL_HAS_CUDA_COMPILER() && _CCCL_CUDACC_BELOW(12) && !defined(CCCL_IGNORE_DEPRECATED_CUDA_BELOW_12)
# error "CUDA versions below 12 are not supported." \
"Define CCCL_IGNORE_DEPRECATED_CUDA_BELOW_12 to suppress this message."
#endif
Expand Down
4 changes: 2 additions & 2 deletions libcudacxx/include/cuda/std/__cccl/diagnostic.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -90,7 +90,7 @@
#endif // !_CCCL_COMPILER(CLANG) && !_CCCL_COMPILER(GCC) && !_CCCL_COMPILER(NVHPC) && !_CCCL_COMPILER(MSVC)

// Enable us to selectively silence cuda compiler warnings
#if _CCCL_HAS_CUDA_COMPILER
#if _CCCL_HAS_CUDA_COMPILER()
# if _CCCL_CUDA_COMPILER(CLANG)
# define _CCCL_NV_DIAG_SUPPRESS(_WARNING)
# define _CCCL_NV_DIAG_DEFAULT(_WARNING)
Expand All @@ -114,7 +114,7 @@
# define _CCCL_NV_DIAG_DEFAULT(_WARNING) _CCCL_PRAGMA(diag_default _WARNING)
# endif // !_CCCL_COMPILER(GCC)
# endif // !__NVCC_DIAG_PRAGMA_SUPPORT__
#else // ^^^ _CCCL_HAS_CUDA_COMPILER ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER vvv
#else // ^^^ _CCCL_HAS_CUDA_COMPILER() ^^^ / vvv !_CCCL_HAS_CUDA_COMPILER() vvv
# define _CCCL_NV_DIAG_SUPPRESS(_WARNING)
# define _CCCL_NV_DIAG_DEFAULT(_WARNING)
#endif // other compilers
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/std/__cccl/execution_space.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@
#endif // no system header

// We need to ensure that we not only compile with a cuda compiler but also compile cuda source files
#if _CCCL_HAS_CUDA_COMPILER && (defined(__CUDACC__) || defined(_NVHPC_CUDA))
#if _CCCL_HAS_CUDA_COMPILER() && (defined(__CUDACC__) || defined(_NVHPC_CUDA))
# define _CCCL_HOST __host__
# define _CCCL_DEVICE __device__
# define _CCCL_HOST_DEVICE __host__ __device__
Expand Down
6 changes: 3 additions & 3 deletions libcudacxx/include/cuda/std/__cccl/extended_data_types.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -51,12 +51,12 @@
#endif

// FIXME: Enable this for clang-cuda in a followup
#if !_CCCL_HAS_CUDA_COMPILER
#if !_CCCL_HAS_CUDA_COMPILER()
# undef _CCCL_HAS_LONG_DOUBLE
# define _CCCL_HAS_LONG_DOUBLE() 1
#endif // !_CCCL_HAS_CUDA_COMPILER
#endif // !_CCCL_HAS_CUDA_COMPILER()

#if _CCCL_HAS_INCLUDE(<cuda_fp16.h>) && (_CCCL_HAS_CUDA_COMPILER || defined(LIBCUDACXX_ENABLE_HOST_NVFP16)) \
#if _CCCL_HAS_INCLUDE(<cuda_fp16.h>) && (_CCCL_HAS_CUDA_COMPILER() || defined(LIBCUDACXX_ENABLE_HOST_NVFP16)) \
&& !defined(CCCL_DISABLE_FP16_SUPPORT)
# undef _CCCL_HAS_NVFP16
# define _CCCL_HAS_NVFP16() 1
Expand Down
2 changes: 1 addition & 1 deletion libcudacxx/include/cuda/std/__cccl/is_non_narrowing_convertible.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -24,7 +24,7 @@
namespace __cccl_internal
{

#if _CCCL_HAS_CUDA_COMPILER && (defined(__CUDACC__) || defined(_NVHPC_CUDA) || _CCCL_COMPILER(NVRTC))
#if _CCCL_HAS_CUDA_COMPILER() && (defined(__CUDACC__) || defined(_NVHPC_CUDA) || _CCCL_COMPILER(NVRTC))
template <class _Tp>
__host__ __device__ _Tp&& __cccl_declval(int);
template <class _Tp>
Expand Down
Loading
Loading