diff --git a/libcudacxx/include/cuda/std/__type_traits/is_extended_floating_point.h b/libcudacxx/include/cuda/std/__type_traits/is_extended_floating_point.h index b9700a87066..040418f5fe7 100644 --- a/libcudacxx/include/cuda/std/__type_traits/is_extended_floating_point.h +++ b/libcudacxx/include/cuda/std/__type_traits/is_extended_floating_point.h @@ -22,16 +22,16 @@ #include -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) # include -#endif // _LIBCUDACXX_HAS_NVFP16 +#endif // _CCCL_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#if defined(_CCCL_HAS_NVBF16) _CCCL_DIAG_PUSH _CCCL_DIAG_SUPPRESS_CLANG("-Wunused-function") # include _CCCL_DIAG_POP -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() # include @@ -53,7 +53,7 @@ _CCCL_INLINE_VAR constexpr bool __is_extended_floating_point_v # endif // !_CCCL_NO_INLINE_VARIABLES #endif // !_CCCL_NO_VARIABLE_TEMPLATES -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) template <> struct __is_extended_floating_point<__half> : true_type {}; @@ -62,9 +62,9 @@ struct __is_extended_floating_point<__half> : true_type template <> _CCCL_INLINE_VAR constexpr bool __is_extended_floating_point_v<__half> = true; # endif // !_CCCL_NO_INLINE_VARIABLES -#endif // _LIBCUDACXX_HAS_NVFP16 +#endif // _CCCL_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#if defined(_CCCL_HAS_NVBF16) template <> struct __is_extended_floating_point<__nv_bfloat16> : true_type {}; @@ -73,7 +73,7 @@ struct __is_extended_floating_point<__nv_bfloat16> : true_type template <> _CCCL_INLINE_VAR constexpr bool __is_extended_floating_point_v<__nv_bfloat16> = true; # endif // !_CCCL_NO_INLINE_VARIABLES -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() template <> diff --git a/libcudacxx/include/cuda/std/limits b/libcudacxx/include/cuda/std/limits index 2bd0191af43..637b80aa305 100644 --- a/libcudacxx/include/cuda/std/limits +++ b/libcudacxx/include/cuda/std/limits @@ -609,7 +609,13 @@ public: #endif // !_LIBCUDACXX_HAS_NO_LONG_DOUBLE }; -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) +# ifdef _LIBCUDACXX_HAS_NVFP16 +# define _LIBCUDACXX_FP16_CONSTEXPR constexpr +# else //_LIBCUDACXX_HAS_NVFP16 +# define _LIBCUDACXX_FP16_CONSTEXPR +# endif //_LIBCUDACXX_HAS_NVFP16 + template <> class __numeric_limits_impl<__half, __numeric_limits_type::__floating_point> { @@ -622,15 +628,15 @@ public: static constexpr int digits = 11; static constexpr int digits10 = 3; static constexpr int max_digits10 = 5; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type min() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type min() noexcept { return type(__half_raw{0x0400u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type max() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type max() noexcept { return type(__half_raw{0x7bffu}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type lowest() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type lowest() noexcept { return type(__half_raw{0xfbffu}); } @@ -638,11 +644,11 @@ public: static constexpr bool is_integer = false; static constexpr bool is_exact = false; static constexpr int radix = __FLT_RADIX__; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type epsilon() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type epsilon() noexcept { return type(__half_raw{0x1400u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type round_error() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type round_error() noexcept { return type(__half_raw{0x3800u}); } @@ -657,19 +663,19 @@ public: static constexpr bool has_signaling_NaN = true; static constexpr float_denorm_style has_denorm = denorm_present; static constexpr bool has_denorm_loss = false; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type infinity() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type infinity() noexcept { return type(__half_raw{0x7c00u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type quiet_NaN() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type quiet_NaN() noexcept { return type(__half_raw{0x7e00u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type signaling_NaN() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type signaling_NaN() noexcept { return type(__half_raw{0x7d00u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type denorm_min() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_FP16_CONSTEXPR type denorm_min() noexcept { return type(__half_raw{0x0001u}); } @@ -682,9 +688,16 @@ public: static constexpr bool tinyness_before = false; static constexpr float_round_style round_style = round_to_nearest; }; -#endif // _LIBCUDACXX_HAS_NVFP16 +# undef _LIBCUDACXX_FP16_CONSTEXPR +#endif // _CCCL_HAS_NVFP16 + +#if defined(_CCCL_HAS_NVBF16) +# ifdef _LIBCUDACXX_HAS_NVBF16 +# define _LIBCUDACXX_BF16_CONSTEXPR constexpr +# else //_LIBCUDACXX_HAS_NVBF16 +# define _LIBCUDACXX_BF16_CONSTEXPR +# endif //_LIBCUDACXX_HAS_NVBF16 -#if defined(_LIBCUDACXX_HAS_NVBF16) template <> class __numeric_limits_impl<__nv_bfloat16, __numeric_limits_type::__floating_point> { @@ -697,15 +710,15 @@ public: static constexpr int digits = 8; static constexpr int digits10 = 2; static constexpr int max_digits10 = 4; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type min() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type min() noexcept { return type(__nv_bfloat16_raw{0x0080u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type max() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type max() noexcept { return type(__nv_bfloat16_raw{0x7f7fu}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type lowest() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type lowest() noexcept { return type(__nv_bfloat16_raw{0xff7fu}); } @@ -713,11 +726,11 @@ public: static constexpr bool is_integer = false; static constexpr bool is_exact = false; static constexpr int radix = __FLT_RADIX__; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type epsilon() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type epsilon() noexcept { return type(__nv_bfloat16_raw{0x3c00u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type round_error() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type round_error() noexcept { return type(__nv_bfloat16_raw{0x3f00u}); } @@ -732,19 +745,19 @@ public: static constexpr bool has_signaling_NaN = true; static constexpr float_denorm_style has_denorm = denorm_present; static constexpr bool has_denorm_loss = false; - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type infinity() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type infinity() noexcept { return type(__nv_bfloat16_raw{0x7f80u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type quiet_NaN() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type quiet_NaN() noexcept { return type(__nv_bfloat16_raw{0x7fc0u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type signaling_NaN() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type signaling_NaN() noexcept { return type(__nv_bfloat16_raw{0x7fa0u}); } - _LIBCUDACXX_HIDE_FROM_ABI static constexpr type denorm_min() noexcept + _LIBCUDACXX_HIDE_FROM_ABI static _LIBCUDACXX_BF16_CONSTEXPR type denorm_min() noexcept { return type(__nv_bfloat16_raw{0x0001u}); } @@ -757,7 +770,8 @@ public: static constexpr bool tinyness_before = false; static constexpr float_round_style round_style = round_to_nearest; }; -#endif // _LIBCUDACXX_HAS_NVBF16 +# undef _LIBCUDACXX_BF16_CONSTEXPR +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() # if defined(_CCCL_BUILTIN_BIT_CAST) || _CCCL_STD_VER >= 2014 diff --git a/libcudacxx/test/libcudacxx/libcxx/utilities/meta/meta.unary/meta.unary.cat/is_floating_point.pass.cpp b/libcudacxx/test/libcudacxx/libcxx/utilities/meta/meta.unary/meta.unary.cat/is_floating_point.pass.cpp index b0b7a3f3b69..5a04070c598 100644 --- a/libcudacxx/test/libcudacxx/libcxx/utilities/meta/meta.unary/meta.unary.cat/is_floating_point.pass.cpp +++ b/libcudacxx/test/libcudacxx/libcxx/utilities/meta/meta.unary/meta.unary.cat/is_floating_point.pass.cpp @@ -80,12 +80,12 @@ int main(int, char**) test_is_floating_point(); test_is_floating_point(); test_is_floating_point(); -#ifdef _LIBCUDACXX_HAS_NVFP16 +#ifdef _CCCL_HAS_NVFP16 test_is_floating_point<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#ifdef _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVFP16 +#ifdef _CCCL_HAS_NVBF16 test_is_floating_point<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test_is_floating_point<__nv_fp8_e4m3>(); test_is_floating_point<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/is_specialized.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/is_specialized.pass.cpp index 7113c0e2772..adb30091033 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/is_specialized.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/is_specialized.pass.cpp @@ -68,12 +68,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 static_assert(!cuda::std::numeric_limits>::is_specialized, "!cuda::std::numeric_limits >::is_specialized"); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/common.h b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/common.h index 8400071611c..7d15f2ba6b6 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/common.h +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/common.h @@ -17,6 +17,7 @@ #define __CUDA_NO_BFLOAT16_CONVERSIONS__ 1 #define __CUDA_NO_BFLOAT16_OPERATORS__ 1 +#include #include template @@ -42,27 +43,43 @@ __host__ __device__ inline __nv_fp8_e5m2 make_fp8_e5m2(double x, __nv_saturation __host__ __device__ inline bool float_eq(__nv_fp8_e4m3 x, __nv_fp8_e4m3 y) { +# if _CCCL_CUDACC_AT_LEAST(12, 2) return float_eq(__half{__nv_cvt_fp8_to_halfraw(x.__x, __NV_E4M3)}, __half{__nv_cvt_fp8_to_halfraw(y.__x, __NV_E4M3)}); +# else + return ::cuda::std::bit_cast(x) == ::cuda::std::bit_cast(y); +# endif } __host__ __device__ inline bool float_eq(__nv_fp8_e5m2 x, __nv_fp8_e5m2 y) { +# if _CCCL_CUDACC_AT_LEAST(12, 2) return float_eq(__half{__nv_cvt_fp8_to_halfraw(x.__x, __NV_E5M2)}, __half{__nv_cvt_fp8_to_halfraw(y.__x, __NV_E5M2)}); +# else + return ::cuda::std::bit_cast(x) == ::cuda::std::bit_cast(y); +# endif } #endif // _CCCL_HAS_NVFP8 -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) __host__ __device__ inline bool float_eq(__half x, __half y) { +# if _CCCL_CUDACC_AT_LEAST(12, 2) return __heq(x, y); +# else + return __half2float(x) == __half2float(y); +# endif } -#endif // _LIBCUDACXX_HAS_NVFP16 +#endif // _CCCL_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#if defined(_CCCL_HAS_NVBF16) __host__ __device__ inline bool float_eq(__nv_bfloat16 x, __nv_bfloat16 y) { +# if _CCCL_CUDACC_AT_LEAST(12, 2) return __heq(x, y); +# else + return __bfloat162float(x) == __bfloat162float(y); +# endif } -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #endif // NUMERIC_LIMITS_MEMBERS_COMMON_H diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/const_data_members.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/const_data_members.pass.cpp index b095d63afcd..093b5d331be 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/const_data_members.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/const_data_members.pass.cpp @@ -110,12 +110,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test_type(); #endif // _LIBCUDACXX_HAS_NO_LONG_DOUBLE -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test_type<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test_type<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test_type<__nv_fp8_e4m3>(); test_type<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/denorm_min.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/denorm_min.pass.cpp index 475f41a3388..9ea232eaad6 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/denorm_min.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/denorm_min.pass.cpp @@ -66,12 +66,12 @@ int main(int, char**) test(LDBL_TRUE_MIN); # endif #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(5.9604644775390625e-08)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(9.18354961579912115600575419705e-41)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(0.001953125)); test<__nv_fp8_e5m2>(make_fp8_e5m2(0.0000152587890625)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits.pass.cpp index 0d3c910b672..01f6b05543b 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits.pass.cpp @@ -55,12 +55,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, 11>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, 8>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, 3>(); test<__nv_fp8_e5m2, 2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits10.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits10.pass.cpp index bd66aeecfeb..24c53725738 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits10.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/digits10.pass.cpp @@ -74,12 +74,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(); test<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/epsilon.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/epsilon.pass.cpp index 15366bdf308..bb65847df33 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/epsilon.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/epsilon.pass.cpp @@ -57,12 +57,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(LDBL_EPSILON); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(0.0009765625)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(0.0078125)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(0.125)); test<__nv_fp8_e5m2>(make_fp8_e5m2(0.25)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm.pass.cpp index 8fa506b93ce..8d9881580bf 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, cuda::std::denorm_present>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, cuda::std::denorm_present>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, cuda::std::denorm_present>(); test<__nv_fp8_e5m2, cuda::std::denorm_present>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm_loss.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm_loss.pass.cpp index 3b7722acd8b..5a046a9b339 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm_loss.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_denorm_loss.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_infinity.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_infinity.pass.cpp index ebddcb4421e..768e53d1c88 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_infinity.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_infinity.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, true>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_quiet_NaN.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_quiet_NaN.pass.cpp index 908f2d7fa4a..4c3e11a9b05 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_quiet_NaN.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_quiet_NaN.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, true>(); test<__nv_fp8_e5m2, true>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_signaling_NaN.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_signaling_NaN.pass.cpp index 62d81c8a524..1b80d1869e6 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_signaling_NaN.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/has_signaling_NaN.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, true>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/infinity.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/infinity.pass.cpp index 627105a4a8c..8dd611556c5 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/infinity.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/infinity.pass.cpp @@ -64,12 +64,12 @@ int main(int, char**) # ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(1. / 0.); # endif -# if defined(_LIBCUDACXX_HAS_NVFP16) +# if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(1.0 / 0.0)); -# endif // _LIBCUDACXX_HAS_NVFP16 -# if defined(_LIBCUDACXX_HAS_NVBF16) +# endif // _CCCL_HAS_NVFP16 +# if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(1.0 / 0.0)); -# endif // _LIBCUDACXX_HAS_NVBF16 +# endif // _CCCL_HAS_NVBF16 # if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(__nv_fp8_e4m3{}); test<__nv_fp8_e5m2>(make_fp8_e5m2(1.0 / 0.0)); @@ -81,12 +81,12 @@ int main(int, char**) # ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(INFINITY); # endif -# if defined(_LIBCUDACXX_HAS_NVFP16) +# if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(INFINITY)); -# endif // _LIBCUDACXX_HAS_NVFP16 -# if defined(_LIBCUDACXX_HAS_NVBF16) +# endif // _CCCL_HAS_NVFP16 +# if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(INFINITY)); -# endif // _LIBCUDACXX_HAS_NVBF16 +# endif // _CCCL_HAS_NVBF16 # if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(__nv_fp8_e4m3{}); test<__nv_fp8_e5m2>(make_fp8_e5m2(INFINITY)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_bounded.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_bounded.pass.cpp index eeb9740e4e2..e28ab8313b6 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_bounded.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_bounded.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, true>(); test<__nv_fp8_e5m2, true>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_exact.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_exact.pass.cpp index c3c2e027c72..e6038f1589b 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_exact.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_exact.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_iec559.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_iec559.pass.cpp index 7bab40e8826..1ff809bad09 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_iec559.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_iec559.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif // _LIBCUDACXX_HAS_NO_LONG_DOUBLE -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_integer.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_integer.pass.cpp index 68e7437f1e0..eed9d38c050 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_integer.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_integer.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_modulo.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_modulo.pass.cpp index 992be2b18b7..fc3ca9dbb4e 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_modulo.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_modulo.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_signed.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_signed.pass.cpp index be7e4f235a7..54005f6c0b9 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_signed.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/is_signed.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, true>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, true>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, true>(); test<__nv_fp8_e5m2, true>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/lowest.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/lowest.pass.cpp index 6a8b2a9c181..72190bd2ad7 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/lowest.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/lowest.pass.cpp @@ -66,12 +66,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(-LDBL_MAX); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(-65504.0)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(-3.3895313892515355e+38)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(-448.0)); test<__nv_fp8_e5m2>(make_fp8_e5m2(-57344.0)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max.pass.cpp index a1582e41b22..5039f773a2f 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max.pass.cpp @@ -65,12 +65,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(LDBL_MAX); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(65504.0)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(3.3895313892515355e+38)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(448.0)); test<__nv_fp8_e5m2>(make_fp8_e5m2(57344.0)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_digits10.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_digits10.pass.cpp index d01a4aa099c..309279bc79c 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_digits10.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_digits10.pass.cpp @@ -69,12 +69,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(); test<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent.pass.cpp index 3027e9f06f5..606e9c52b7f 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent.pass.cpp @@ -62,12 +62,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, 16>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, 128>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, 8>(); test<__nv_fp8_e5m2, 15>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent10.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent10.pass.cpp index 5924aee173d..61145deec86 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent10.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/max_exponent10.pass.cpp @@ -62,12 +62,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, 4>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, 38>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, 2>(); test<__nv_fp8_e5m2, 4>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min.pass.cpp index 15f470909df..ccab08a38f5 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min.pass.cpp @@ -66,12 +66,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(LDBL_MIN); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(6.103515625e-05)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(1.17549435082228750796873653722e-38)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(0.015625)); test<__nv_fp8_e5m2>(make_fp8_e5m2(0.000061035)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent.pass.cpp index b63d653a7c3..c942a6288be 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent.pass.cpp @@ -62,12 +62,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, -13>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, -125>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, -6>(); test<__nv_fp8_e5m2, -15>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent10.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent10.pass.cpp index a6ff20e7fde..e9b6f29d25f 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent10.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/min_exponent10.pass.cpp @@ -62,12 +62,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, -4>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, -37>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, -2>(); test<__nv_fp8_e5m2, -5>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/quiet_NaN.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/quiet_NaN.pass.cpp index 2d6d9582f5c..a8b076fbeee 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/quiet_NaN.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/quiet_NaN.pass.cpp @@ -108,12 +108,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(); test<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/radix.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/radix.pass.cpp index 7e5c87927aa..dd15c391180 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/radix.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/radix.pass.cpp @@ -55,12 +55,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, FLT_RADIX>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, FLT_RADIX>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, FLT_RADIX>(); test<__nv_fp8_e5m2, FLT_RADIX>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_error.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_error.pass.cpp index d4faf373a09..95ed80eb951 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_error.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_error.pass.cpp @@ -57,12 +57,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(0.5); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(__double2half(0.5)); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(__double2bfloat16(0.5)); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(make_fp8_e4m3(0.5)); test<__nv_fp8_e5m2>(make_fp8_e5m2(0.5)); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_style.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_style.pass.cpp index 8515581d650..1eb5c0b0f5a 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_style.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/round_style.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, cuda::std::round_to_nearest>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, cuda::std::round_to_nearest>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, cuda::std::round_to_nearest>(); test<__nv_fp8_e5m2, cuda::std::round_to_nearest>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/signaling_NaN.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/signaling_NaN.pass.cpp index 19ace1b3d2c..0ec70976b32 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/signaling_NaN.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/signaling_NaN.pass.cpp @@ -108,12 +108,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3>(); test<__nv_fp8_e5m2>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/tinyness_before.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/tinyness_before.pass.cpp index 38dec8c872b..1da28874b06 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/tinyness_before.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/tinyness_before.pass.cpp @@ -54,12 +54,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/traps.pass.cpp b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/traps.pass.cpp index 55d7eb990db..4cb627a4b77 100644 --- a/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/traps.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/language.support/support.limits/limits/numeric.limits.members/traps.pass.cpp @@ -60,12 +60,12 @@ int main(int, char**) #ifndef _LIBCUDACXX_HAS_NO_LONG_DOUBLE test(); #endif -#if defined(_LIBCUDACXX_HAS_NVFP16) +#if defined(_CCCL_HAS_NVFP16) test<__half, false>(); -#endif // _LIBCUDACXX_HAS_NVFP16 -#if defined(_LIBCUDACXX_HAS_NVBF16) +#endif // _CCCL_HAS_NVFP16 +#if defined(_CCCL_HAS_NVBF16) test<__nv_bfloat16, false>(); -#endif // _LIBCUDACXX_HAS_NVBF16 +#endif // _CCCL_HAS_NVBF16 #if _CCCL_HAS_NVFP8() test<__nv_fp8_e4m3, false>(); test<__nv_fp8_e5m2, false>(); diff --git a/libcudacxx/test/libcudacxx/std/utilities/template.bitset/bitset.members/all.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/template.bitset/bitset.members/all.pass.cpp index 76bc4de94df..5cdbc309873 100644 --- a/libcudacxx/test/libcudacxx/std/utilities/template.bitset/bitset.members/all.pass.cpp +++ b/libcudacxx/test/libcudacxx/std/utilities/template.bitset/bitset.members/all.pass.cpp @@ -47,7 +47,8 @@ __host__ __device__ TEST_CONSTEXPR_CXX14 bool test() int main(int, char**) { test(); -#if TEST_STD_VER >= 2014 +// this test spuriously fails for nvcc 11.1 +#if TEST_STD_VER >= 2014 && !_CCCL_CUDA_COMPILER(NVCC, <, 12) static_assert(test(), ""); #endif // TEST_STD_VER >= 2014