Skip to content

Commit f9835a5

Browse files
davebayerdavebayer
authored
Replace _CCCL_HAS_CUDA_COMPILER() with _CCCL_CUDA_COMPILATION() (#6399)
1 parent 4f60db5 commit f9835a5

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

47 files changed

+106
-106
lines changed

‎cub/cub/util_device.cuh‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -754,6 +754,6 @@ private:
754754
};
755755
CUB_NAMESPACE_END
756756

757-
#if _CCCL_HAS_CUDA_COMPILER() && !_CCCL_COMPILER(NVRTC)
757+
#if _CCCL_CUDA_COMPILATION() && !_CCCL_COMPILER(NVRTC)
758758
# include <cub/detail/launcher/cuda_runtime.cuh> // to complete the definition of TripleChevronFactory
759-
#endif // _CCCL_HAS_CUDA_COMPILER() && !_CCCL_COMPILER(NVRTC)
759+
#endif // _CCCL_CUDA_COMPILATION() && !_CCCL_COMPILER(NVRTC)

‎cudax/include/cuda/experimental/__container/async_buffer.cuh‎

Lines changed: 7 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -21,9 +21,9 @@
2121
# pragma system_header
2222
#endif // no system header
2323

24-
#if _CCCL_HAS_CUDA_COMPILER()
24+
#if _CCCL_CUDA_COMPILATION()
2525
# include <cub/device/device_transform.cuh>
26-
#endif // _CCCL_HAS_CUDA_COMPILER()
26+
#endif // _CCCL_CUDA_COMPILATION()
2727

2828
#include <cuda/__container/heterogeneous_iterator.h>
2929
#include <cuda/__container/uninitialized_async_buffer.h>
@@ -691,14 +691,13 @@ __fill_n(cuda::stream_ref __stream, _Tp* __first, ::cuda::std::size_t __count, c
691691
}
692692
else
693693
{
694-
#if _CCCL_HAS_CUDA_COMPILER()
694+
#if _CCCL_CUDA_COMPILATION()
695695
::cuda::experimental::__ensure_current_device __guard(__stream);
696696
::cub::DeviceTransform::Fill(__first, __count, __value, __stream.get());
697-
#else
698-
static_assert(0,
699-
"CUDA compiler is required to initialize a buffer with "
700-
"elements larger than 4 bytes");
701-
#endif
697+
#else // ^^^ _CCCL_CUDA_COMPILATION() ^^^ / vvv !_CCCL_CUDA_COMPILATION() vvv
698+
static_assert(sizeof(_Tp) <= 4,
699+
"CUDA compiler is required to initialize an async_buffer with elements larger than 4 bytes");
700+
#endif // ^^^ !_CCCL_CUDA_COMPILATION() ^^^
702701
}
703702
}
704703
}

‎libcudacxx/codegen/generators/header.h‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -58,7 +58,7 @@ inline void FormatHeader(std::ostream& out)
5858
5959
_CCCL_BEGIN_NAMESPACE_CUDA_STD
6060
61-
#if _CCCL_HAS_CUDA_COMPILER()
61+
#if _CCCL_CUDA_COMPILATION()
6262
6363
extern "C" _CCCL_DEVICE void __atomic_cas_128b_unsupported_before_SM_90();
6464
extern "C" _CCCL_DEVICE void __atomic_exchange_128b_unsupported_before_SM_90();
@@ -71,7 +71,7 @@ extern "C" _CCCL_DEVICE void __atomic_ldst_128b_unsupported_before_SM_70();
7171
inline void FormatTail(std::ostream& out)
7272
{
7373
std::string tail = R"XXX(
74-
#endif // _CCCL_HAS_CUDA_COMPILER()
74+
#endif // _CCCL_CUDA_COMPILATION()
7575
7676
_CCCL_END_NAMESPACE_CUDA_STD
7777

‎libcudacxx/include/cuda/__memcpy_async/memcpy_async.h‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -22,7 +22,7 @@
2222
# pragma system_header
2323
#endif // no system header
2424

25-
#if _CCCL_HAS_CUDA_COMPILER()
25+
#if _CCCL_CUDA_COMPILATION()
2626

2727
# include <cuda/__barrier/async_contract_fulfillment.h>
2828
# include <cuda/__barrier/barrier.h>
@@ -174,6 +174,6 @@ _CCCL_END_NAMESPACE_CUDA
174174

175175
# include <cuda/std/__cccl/epilogue.h>
176176

177-
#endif // _CCCL_HAS_CUDA_COMPILER()
177+
#endif // _CCCL_CUDA_COMPILATION()
178178

179179
#endif // _CUDA___MEMCPY_ASYNC_MEMCPY_ASYNC_H_

‎libcudacxx/include/cuda/std/__atomic/functions/cuda_ptx_generated.h‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -41,7 +41,7 @@
4141

4242
_CCCL_BEGIN_NAMESPACE_CUDA_STD
4343

44-
#if _CCCL_HAS_CUDA_COMPILER()
44+
#if _CCCL_CUDA_COMPILATION()
4545

4646
extern "C" _CCCL_DEVICE void __atomic_cas_128b_unsupported_before_SM_90();
4747
extern "C" _CCCL_DEVICE void __atomic_exchange_128b_unsupported_before_SM_90();
@@ -4425,7 +4425,7 @@ template <class _Type, class _Up, class _Sco>
44254425
return __atomic_fetch_add_cuda(__ptr, -__op, __memorder, _Sco{});
44264426
}
44274427

4428-
#endif // _CCCL_HAS_CUDA_COMPILER()
4428+
#endif // _CCCL_CUDA_COMPILATION()
44294429

44304430
_CCCL_END_NAMESPACE_CUDA_STD
44314431

‎thrust/examples/include/host_device.h‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -18,7 +18,7 @@
1818

1919
#include <cuda/__cccl_config>
2020

21-
#if !_CCCL_HAS_CUDA_COMPILER()
21+
#if !_CCCL_CUDA_COMPILATION()
2222

2323
# ifndef __host__
2424
# define __host__

‎thrust/thrust/random/detail/normal_distribution_base.h‎

Lines changed: 4 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -146,11 +146,12 @@ class normal_distribution_portable
146146
template <typename RealType>
147147
struct normal_distribution_base
148148
{
149-
#if _CCCL_HAS_CUDA_COMPILER() && !_CCCL_CUDA_COMPILER(NVHPC)
149+
#if _CCCL_CUDA_COMPILATION() && !_CCCL_CUDA_COMPILER(NVHPC)
150150
using type = normal_distribution_nvcc<RealType>;
151-
#else
151+
#else // ^^^ _CCCL_CUDA_COMPILATION() && !_CCCL_CUDA_COMPILER(NVHPC) ^^^ /
152+
// vvv !_CCCL_CUDA_COMPILATION() || _CCCL_CUDA_COMPILER(NVHPC) vvv
152153
using type = normal_distribution_portable<RealType>;
153-
#endif
154+
#endif // ^^^ !_CCCL_CUDA_COMPILATION() || _CCCL_CUDA_COMPILER(NVHPC) ^^^
154155
};
155156
} // namespace random::detail
156157
THRUST_NAMESPACE_END

‎thrust/thrust/system/cuda/detail/adjacent_difference.h‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -36,7 +36,7 @@
3636
# pragma system_header
3737
#endif // no system header
3838

39-
#if _CCCL_HAS_CUDA_COMPILER()
39+
#if _CCCL_CUDA_COMPILATION()
4040

4141
# include <thrust/system/cuda/config.h>
4242

@@ -215,4 +215,4 @@ THRUST_NAMESPACE_END
215215
//
216216
# include <thrust/adjacent_difference.h>
217217
# include <thrust/memory.h>
218-
#endif
218+
#endif // _CCCL_CUDA_COMPILATION()

‎thrust/thrust/system/cuda/detail/assign_value.h‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -26,7 +26,7 @@
2626
# pragma system_header
2727
#endif // no system header
2828

29-
#if _CCCL_HAS_CUDA_COMPILER()
29+
#if _CCCL_CUDA_COMPILATION()
3030

3131
# include <thrust/detail/raw_pointer_cast.h>
3232
# include <thrust/system/cuda/detail/cross_system.h>
@@ -121,4 +121,4 @@ _CCCL_HOST_DEVICE void assign_value(cross_system<System1, System2>& systems, Poi
121121
}
122122
} // namespace cuda_cub
123123
THRUST_NAMESPACE_END
124-
#endif
124+
#endif // _CCCL_CUDA_COMPILATION()

‎thrust/thrust/system/cuda/detail/copy.h‎

Lines changed: 8 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -48,9 +48,9 @@
4848
#include <thrust/system/cuda/detail/util.h>
4949
#include <thrust/type_traits/is_trivially_relocatable.h>
5050

51-
#if _CCCL_HAS_CUDA_COMPILER()
51+
#if _CCCL_CUDA_COMPILATION()
5252
# include <cub/device/dispatch/tuning/tuning_transform.cuh>
53-
#endif // _CCCL_HAS_CUDA_COMPILER()
53+
#endif // _CCCL_CUDA_COMPILATION()
5454

5555
#include <cuda/__fwd/zip_iterator.h>
5656
#include <cuda/std/tuple>
@@ -131,7 +131,7 @@ OutputIt _CCCL_HOST non_trivial_cross_system_copy_n(
131131
return ret;
132132
}
133133

134-
#if _CCCL_HAS_CUDA_COMPILER()
134+
#if _CCCL_CUDA_COMPILATION()
135135
// non-trivial copy D->H, only supported with NVCC compiler
136136
// because copy ctor must have __device__ annotations, which is nvcc-only
137137
// feature
@@ -158,7 +158,7 @@ OutputIt _CCCL_HOST non_trivial_cross_system_copy_n(
158158
OutputIt ret = thrust::copy_n(host_s, temp_host.data(), num_items, result);
159159
return ret;
160160
}
161-
#endif // _CCCL_HAS_CUDA_COMPILER()
161+
#endif // _CCCL_CUDA_COMPILATION()
162162

163163
template <class System1, class System2, class InputIt, class Size, class OutputIt>
164164
OutputIt _CCCL_HOST cross_system_copy_n(cross_system<System1, System2> systems, InputIt begin, Size n, OutputIt result)
@@ -188,7 +188,7 @@ OutputIt _CCCL_HOST cross_system_copy_n(cross_system<System1, System2> systems,
188188
}
189189
}
190190

191-
#if _CCCL_HAS_CUDA_COMPILER()
191+
#if _CCCL_CUDA_COMPILATION()
192192
template <class Derived, class InputIt, class OutputIt>
193193
OutputIt THRUST_RUNTIME_FUNCTION
194194
device_to_device(execution_policy<Derived>& policy, InputIt first, InputIt last, OutputIt result)
@@ -229,10 +229,10 @@ device_to_device(execution_policy<Derived>& policy, InputIt first, InputIt last,
229229
policy, first, last, result, ::cuda::proclaim_copyable_arguments(::cuda::std::identity{}));
230230
}
231231
}
232-
#endif // _CCCL_HAS_CUDA_COMPILER()
232+
#endif // _CCCL_CUDA_COMPILATION()
233233
} // namespace __copy
234234

235-
#if _CCCL_HAS_CUDA_COMPILER()
235+
#if _CCCL_CUDA_COMPILATION()
236236

237237
_CCCL_EXEC_CHECK_DISABLE
238238
template <class System, class InputIterator, class OutputIterator>
@@ -253,7 +253,7 @@ copy_n(execution_policy<System>& system, InputIterator first, Size n, OutputIter
253253
(result = thrust::copy_n(cvt_to_seq(derived_cast(system)), first, n, result);));
254254
return result;
255255
}
256-
#endif // _CCCL_HAS_CUDA_COMPILER()
256+
#endif // _CCCL_CUDA_COMPILATION()
257257

258258
template <class System1, class System2, class InputIterator, class OutputIterator>
259259
OutputIterator _CCCL_HOST

0 commit comments

Comments
 (0)