|
17 | 17 | #include <cub/device/dispatch/dispatch_merge.cuh> |
18 | 18 | #include <cub/util_namespace.cuh> |
19 | 19 |
|
20 | | -#include <cuda/__execution/determinism.h> |
21 | 20 | #include <cuda/__execution/require.h> |
22 | 21 | #include <cuda/std/__functional/operations.h> |
23 | 22 | #include <cuda/std/cstdint> |
@@ -193,15 +192,6 @@ struct DeviceMerge |
193 | 192 | { |
194 | 193 | _CCCL_NVTX_RANGE_SCOPE("cub::DeviceMerge::MergeKeys"); |
195 | 194 |
|
196 | | - using requirements_t = ::cuda::std::execution:: |
197 | | - __query_result_or_t<EnvT, ::cuda::execution::__get_requirements_t, ::cuda::std::execution::env<>>; |
198 | | - using requested_determinism_t = |
199 | | - ::cuda::std::execution::__query_result_or_t<requirements_t, |
200 | | - ::cuda::execution::determinism::__get_determinism_t, |
201 | | - ::cuda::execution::determinism::run_to_run_t>; |
202 | | - static_assert(!::cuda::std::is_same_v<requested_determinism_t, ::cuda::execution::determinism::gpu_to_gpu_t>, |
203 | | - "gpu_to_gpu determinism is not supported for unstable device merge"); |
204 | | - |
205 | 195 | return detail::dispatch_with_env( |
206 | 196 | env, [&]([[maybe_unused]] auto tuning, void* d_temp_storage, size_t& temp_storage_bytes, cudaStream_t stream) { |
207 | 197 | return detail::merge::dispatch( |
@@ -423,15 +413,6 @@ struct DeviceMerge |
423 | 413 | { |
424 | 414 | _CCCL_NVTX_RANGE_SCOPE("cub::DeviceMerge::MergePairs"); |
425 | 415 |
|
426 | | - using requirements_t = ::cuda::std::execution:: |
427 | | - __query_result_or_t<EnvT, ::cuda::execution::__get_requirements_t, ::cuda::std::execution::env<>>; |
428 | | - using requested_determinism_t = |
429 | | - ::cuda::std::execution::__query_result_or_t<requirements_t, |
430 | | - ::cuda::execution::determinism::__get_determinism_t, |
431 | | - ::cuda::execution::determinism::run_to_run_t>; |
432 | | - static_assert(!::cuda::std::is_same_v<requested_determinism_t, ::cuda::execution::determinism::gpu_to_gpu_t>, |
433 | | - "gpu_to_gpu determinism is not supported for unstable device merge"); |
434 | | - |
435 | 416 | return detail::dispatch_with_env( |
436 | 417 | env, [&]([[maybe_unused]] auto tuning, void* d_temp_storage, size_t& temp_storage_bytes, cudaStream_t stream) { |
437 | 418 | return detail::merge::dispatch( |
|
0 commit comments