From 19fb1f418ee2b0f4008ba5c7b17f45ebe9443dc5 Mon Sep 17 00:00:00 2001 From: TriveniTadapaneni Date: Mon, 27 Jul 2026 17:34:56 +0530 Subject: [PATCH] Revert "feat(rocthrust): CCCL 3.0.x changes (#3773)" This reverts commit 1eb022d06b7e2a1b0031d74e96d85a2e9e150c99. --- .../rocprim/test_device_segmented_topk.cpp | 40 - .../rocprim/test/rocprim/test_device_topk.cpp | 40 - projects/rocthrust/.clang-format | 2 +- projects/rocthrust/CHANGELOG.md | 9 - projects/rocthrust/CMakeLists.txt | 17 +- .../bench/adjacent_difference/basic.cu | 9 +- .../bench/adjacent_difference/custom.cu | 9 +- .../bench/adjacent_difference/in_place.cu | 9 +- .../rocthrust/benchmark/bench/copy/basic.cu | 9 +- projects/rocthrust/benchmark/bench/copy/if.cu | 14 +- .../rocthrust/benchmark/bench/equal/basic.cu | 9 +- .../rocthrust/benchmark/bench/fill/basic.cu | 9 +- .../benchmark/bench/for_each/basic.cu | 9 +- .../benchmark/bench/inner_product/basic.cu | 9 +- .../rocthrust/benchmark/bench/merge/basic.cu | 9 +- .../benchmark/bench/partition/basic.cu | 9 +- .../rocthrust/benchmark/bench/reduce/basic.cu | 9 +- .../benchmark/bench/reduce/by_key.cu | 14 +- .../benchmark/bench/scan/exclusive/by_key.cu | 9 +- .../benchmark/bench/scan/exclusive/max.cu | 9 +- .../benchmark/bench/scan/exclusive/sum.cu | 9 +- .../benchmark/bench/scan/inclusive/by_key.cu | 9 +- .../benchmark/bench/scan/inclusive/max.cu | 9 +- .../benchmark/bench/scan/inclusive/sum.cu | 9 +- .../benchmark/bench/set_operations/base.hpp | 6 +- .../benchmark/bench/set_operations/by_key.hpp | 6 +- .../bench/set_operations/difference.cu | 4 +- .../bench/set_operations/difference_by_key.cu | 4 +- .../bench/set_operations/intersection.cu | 4 +- .../set_operations/intersection_by_key.cu | 4 +- .../set_operations/symmetric_difference.cu | 4 +- .../symmetric_difference_by_key.cu | 4 +- .../benchmark/bench/set_operations/union.cu | 4 +- .../bench/set_operations/union_by_key.cu | 4 +- .../benchmark/bench/shuffle/basic.cu | 9 +- .../rocthrust/benchmark/bench/sort/keys.cu | 9 +- .../benchmark/bench/sort/keys_custom.cu | 11 +- .../rocthrust/benchmark/bench/sort/pairs.cu | 9 +- .../benchmark/bench/sort/pairs_custom.cu | 9 +- .../benchmark/bench/tabulate/basic.cu | 9 +- .../benchmark/bench/transform/basic.cu | 28 +- .../benchmark/bench/transform_reduce/sum.cu | 9 +- .../rocthrust/benchmark/bench/unique/basic.cu | 9 +- .../benchmark/bench/unique/by_key.cu | 12 +- .../bench/vectorized_search/basic.cu | 7 +- .../bench/vectorized_search/lower_bound.cu | 7 +- .../bench/vectorized_search/upper_bound.cu | 7 +- .../benchmark/bench_utils/bench_utils.hpp | 8 +- projects/rocthrust/cmake/Dependencies.cmake | 6 +- projects/rocthrust/cmake/FindROCMCmake.cmake | 4 +- projects/rocthrust/docs/doxygen/Doxyfile | 11 +- .../examples/arbitrary_transformation.cu | 35 +- .../rocthrust/examples/counting_iterator.cu | 3 +- .../cuda/custom_temporary_allocation.cu | 80 +- .../rocthrust/examples/cuda/range_view.cu | 12 +- projects/rocthrust/examples/device_ptr.cu | 7 +- projects/rocthrust/examples/expand.cu | 8 +- .../rocthrust/examples/include/host_device.h | 4 +- .../examples/minimal_custom_backend.cu | 12 +- projects/rocthrust/examples/repeated_range.cu | 16 +- projects/rocthrust/examples/set_operations.cu | 5 +- projects/rocthrust/examples/strided_range.cu | 16 +- projects/rocthrust/examples/tiled_range.cu | 16 +- projects/rocthrust/extra/CMakeLists.txt | 6 +- projects/rocthrust/test/CMakeLists.txt | 16 +- .../rocthrust/test/test_address_stability.cpp | 20 +- projects/rocthrust/test/test_allocator.cpp | 5 +- .../test/test_allocator_aware_policies.cpp | 14 +- projects/rocthrust/test/test_async_copy.cpp | 422 +++ .../rocthrust/test/test_async_for_each.cpp | 97 + projects/rocthrust/test/test_async_reduce.cpp | 1019 ++++++ .../rocthrust/test/test_async_reduce_into.cpp | 573 ++++ projects/rocthrust/test/test_async_scan.cpp | 227 ++ projects/rocthrust/test/test_async_sort.cpp | 314 ++ .../rocthrust/test/test_async_transform.cpp | 513 +++ .../rocthrust/test/test_binary_search.cpp | 8 +- projects/rocthrust/test/test_complex.cpp | 22 +- .../rocthrust/test/test_constant_iterator.cpp | 41 +- projects/rocthrust/test/test_copy.cpp | 25 +- .../rocthrust/test/test_counting_iterator.cpp | 65 +- .../test/test_dependencies_aware_policies.cpp | 143 + projects/rocthrust/test/test_dereference.cpp | 3 +- .../rocthrust/test/test_device_delete.cpp | 6 +- .../rocthrust/test/test_discard_iterator.cpp | 78 +- projects/rocthrust/test/test_event.cpp | 181 ++ projects/rocthrust/test/test_find.cpp | 11 +- projects/rocthrust/test/test_functional.cpp | 56 +- projects/rocthrust/test/test_future.cpp | 221 ++ .../test/test_is_contiguous_iterator.cpp | 134 +- .../test/test_is_operator_function_object.cpp | 120 +- .../rocthrust/test/test_is_partitioned.cpp | 3 +- projects/rocthrust/test/test_logical.cpp | 3 +- projects/rocthrust/test/test_memory.cpp | 167 +- projects/rocthrust/test/test_merge_by_key.cpp | 6 +- .../rocthrust/test/test_merge_key_value.cpp | 6 +- projects/rocthrust/test/test_min_and_max.cpp | 66 +- .../rocthrust/test/test_offset_iterator.cpp | 278 -- projects/rocthrust/test/test_pair.cpp | 4 +- projects/rocthrust/test/test_partition.cpp | 6 +- .../rocthrust/test/test_partition_point.cpp | 3 +- .../test/test_permutation_iterator.cpp | 46 +- projects/rocthrust/test/test_random.cpp | 4 +- .../test/test_raw_reference_cast.cpp | 86 - projects/rocthrust/test/test_remove.cpp | 48 +- projects/rocthrust/test/test_replace.cpp | 14 +- projects/rocthrust/test/test_reverse.cpp | 10 +- .../rocthrust/test/test_reverse_iterator.cpp | 38 +- projects/rocthrust/test/test_scan.cpp | 62 +- .../rocthrust/test/test_set_difference.cpp | 4 +- .../rocthrust/test/test_set_intersection.cpp | 4 +- .../test/test_sort_permutation_iterator.cpp | 4 +- projects/rocthrust/test/test_tabulate.cpp | 59 +- .../test/test_tabulate_output_iterator.cpp | 43 +- projects/rocthrust/test/test_transform.cpp | 20 +- .../test_transform_input_output_iterator.cpp | 46 +- .../test/test_transform_iterator.cpp | 127 +- .../test/test_transform_output_iterator.cpp | 41 +- .../rocthrust/test/test_transform_scan.cpp | 3 +- .../rocthrust/test/test_trivial_sequence.cpp | 8 +- projects/rocthrust/test/test_tuple.cpp | 132 +- .../rocthrust/test/test_tuple_algorithms.cpp | 77 + projects/rocthrust/test/test_tuple_reduce.cpp | 4 +- projects/rocthrust/test/test_tuple_scan.cpp | 4 +- projects/rocthrust/test/test_tuple_sort.cpp | 18 +- .../rocthrust/test/test_tuple_transform.cpp | 8 +- projects/rocthrust/test/test_type_traits.cpp | 47 +- .../test/test_uninitialized_copy.cpp | 9 +- .../test/test_uninitialized_fill.cpp | 9 +- projects/rocthrust/test/test_unique.cpp | 9 +- .../rocthrust/test/test_universal_memory.cpp | 16 +- projects/rocthrust/test/test_utils.hpp | 36 +- .../test/test_vector_manipulation.cpp | 4 +- projects/rocthrust/test/test_zip_function.cpp | 58 +- projects/rocthrust/test/test_zip_iterator.cpp | 91 +- projects/rocthrust/testing/CMakeLists.txt | 17 +- .../rocthrust/testing/address_stability.cu | 20 +- projects/rocthrust/testing/allocator.cu | 5 +- .../testing/allocator_aware_policies.cu | 14 +- .../rocthrust/testing/async/CMakeLists.txt | 2 + .../async/exclusive_scan/CMakeLists.txt | 68 + .../async/exclusive_scan/counting_iterator.cu | 66 + .../async/exclusive_scan/discard_output.cu | 63 + .../async/exclusive_scan/large_indices.cu | 269 ++ .../async/exclusive_scan/large_types.cu | 78 + .../async/exclusive_scan/mixed_types.cu | 143 + .../testing/async/exclusive_scan/mixin.h | 132 + .../testing/async/exclusive_scan/simple.cu | 97 + .../async/exclusive_scan/stateful_operator.cu | 88 + .../async/exclusive_scan/using_vs_adl.cu | 196 ++ .../async/inclusive_scan/CMakeLists.txt | 69 + .../async/inclusive_scan/counting_iterator.cu | 65 + .../async/inclusive_scan/discard_output.cu | 61 + .../async/inclusive_scan/large_indices.cu | 263 ++ .../async/inclusive_scan/large_types.cu | 78 + .../async/inclusive_scan/mixed_types.cu | 132 + .../testing/async/inclusive_scan/mixin.h | 148 + .../testing/async/inclusive_scan/simple.cu | 121 + .../async/inclusive_scan/stateful_operator.cu | 88 + .../async/inclusive_scan/using_vs_adl.cu | 194 ++ projects/rocthrust/testing/async/mixin.h | 699 ++++ .../testing/async/test_policy_overloads.h | 371 +++ projects/rocthrust/testing/async_copy.cu | 346 ++ projects/rocthrust/testing/async_for_each.cu | 90 + projects/rocthrust/testing/async_reduce.cu | 928 ++++++ .../rocthrust/testing/async_reduce_into.cu | 512 +++ projects/rocthrust/testing/async_sort.cu | 202 ++ projects/rocthrust/testing/async_transform.cu | 466 +++ projects/rocthrust/testing/binary_search.cu | 8 +- projects/rocthrust/testing/complex.cu | 10 +- .../rocthrust/testing/constant_iterator.cu | 40 +- projects/rocthrust/testing/copy.cu | 25 +- .../rocthrust/testing/counting_iterator.cu | 66 +- .../testing/cuda/adjacent_difference.cu | 6 +- projects/rocthrust/testing/cuda/complex.cu | 10 +- .../rocthrust/testing/cuda/is_partitioned.cu | 4 +- .../rocthrust/testing/cuda/offset_iterator.cu | 137 - projects/rocthrust/testing/cuda/sort.cu | 27 +- .../rocthrust/testing/cuda/stream_wrapper.cu | 104 - projects/rocthrust/testing/cuda/transform.cu | 5 +- .../testing/cuda/transform_iterator.cmake | 12 - .../testing/cuda/transform_iterator.cu | 36 - .../testing/dependencies_aware_policies.cu | 133 + projects/rocthrust/testing/dereference.cu | 3 +- projects/rocthrust/testing/device_delete.cu | 6 +- .../rocthrust/testing/discard_iterator.cu | 51 +- projects/rocthrust/testing/event.cu | 174 + projects/rocthrust/testing/find.cu | 9 +- projects/rocthrust/testing/functional.cu | 54 +- projects/rocthrust/testing/future.cu | 223 ++ .../testing/generate_const_iterators.cu | 46 + .../testing/is_contiguous_iterator.cu | 134 +- .../testing/is_operator_function_object.cu | 122 +- projects/rocthrust/testing/is_partitioned.cu | 3 +- projects/rocthrust/testing/logical.cu | 3 +- projects/rocthrust/testing/memory.cu | 24 +- projects/rocthrust/testing/merge_by_key.cu | 6 +- projects/rocthrust/testing/merge_key_value.cu | 6 +- projects/rocthrust/testing/min_and_max.cu | 66 +- projects/rocthrust/testing/offset_iterator.cu | 244 -- .../rocthrust/testing/omp/reduce_intervals.cu | 6 +- projects/rocthrust/testing/pair.cu | 4 +- projects/rocthrust/testing/partition.cu | 6 +- projects/rocthrust/testing/partition_point.cu | 3 +- .../rocthrust/testing/permutation_iterator.cu | 39 +- .../rocthrust/testing/raw_reference_cast.cu | 82 - projects/rocthrust/testing/remove.cu | 3 +- projects/rocthrust/testing/replace.cu | 14 +- projects/rocthrust/testing/reverse.cu | 6 +- .../rocthrust/testing/reverse_iterator.cu | 37 +- projects/rocthrust/testing/scan.cu | 72 +- projects/rocthrust/testing/set_difference.cu | 4 +- .../rocthrust/testing/set_intersection.cu | 4 +- .../testing/sort_permutation_iterator.cu | 4 +- projects/rocthrust/testing/tabulate.cu | 3 +- .../testing/tabulate_output_iterator.cu | 42 +- projects/rocthrust/testing/transform.cu | 8 +- .../transform_input_output_iterator.cu | 45 +- .../rocthrust/testing/transform_iterator.cu | 75 +- .../testing/transform_output_iterator.cu | 40 +- projects/rocthrust/testing/transform_scan.cu | 3 +- .../rocthrust/testing/trivial_sequence.cu | 8 +- projects/rocthrust/testing/tuple.cu | 133 +- .../rocthrust/testing/tuple_algorithms.cu | 72 + projects/rocthrust/testing/tuple_reduce.cu | 4 +- projects/rocthrust/testing/tuple_scan.cu | 4 +- projects/rocthrust/testing/tuple_sort.cu | 18 +- projects/rocthrust/testing/tuple_transform.cu | 8 +- projects/rocthrust/testing/type_traits.cu | 47 +- .../rocthrust/testing/uninitialized_copy.cu | 9 +- .../rocthrust/testing/uninitialized_fill.cu | 9 +- projects/rocthrust/testing/unique.cu | 9 +- .../rocthrust/testing/unittest/assertions.h | 11 +- projects/rocthrust/testing/unittest/meta.h | 12 +- .../testing/unittest/runtime_static_assert.h | 119 + .../testing/unittest/testframework.h | 23 +- .../rocthrust/testing/unittest/unittest.h | 8 +- projects/rocthrust/testing/unittest/util.h | 6 +- .../rocthrust/testing/unittest/util_async.h | 84 + .../testing/unittest_static_assert.cmake | 15 + .../testing/unittest_static_assert.cu | 50 + .../rocthrust/testing/universal_memory.cu | 13 +- .../rocthrust/testing/vector_manipulation.cu | 4 +- projects/rocthrust/testing/zip_function.cu | 56 +- projects/rocthrust/testing/zip_iterator.cu | 80 +- projects/rocthrust/thrust/addressof.h | 28 +- .../rocthrust/thrust/adjacent_difference.h | 27 +- projects/rocthrust/thrust/advance.h | 16 +- projects/rocthrust/thrust/allocate_unique.h | 128 +- projects/rocthrust/thrust/async/copy.h | 136 + projects/rocthrust/thrust/async/for_each.h | 113 + projects/rocthrust/thrust/async/reduce.h | 331 ++ projects/rocthrust/thrust/async/scan.h | 323 ++ projects/rocthrust/thrust/async/sort.h | 297 ++ projects/rocthrust/thrust/async/transform.h | 148 + projects/rocthrust/thrust/complex.h | 64 +- projects/rocthrust/thrust/copy.h | 25 +- projects/rocthrust/thrust/count.h | 24 +- .../thrust/detail/adjacent_difference.inl | 47 +- projects/rocthrust/thrust/detail/advance.inl | 43 +- .../thrust/detail/algorithm_wrapper.h | 10 - .../detail/allocator/allocator_traits.h | 12 +- .../detail/allocator/allocator_traits.inl | 5 +- .../detail/allocator/copy_construct_range.h | 7 - .../detail/allocator/copy_construct_range.inl | 18 +- .../thrust/detail/allocator/destroy_range.h | 8 - .../thrust/detail/allocator/destroy_range.inl | 4 +- .../detail/allocator/fill_construct_range.h | 8 - .../detail/allocator/malloc_allocator.h | 7 - .../detail/allocator/malloc_allocator.inl | 34 +- .../detail/allocator/no_throw_allocator.h | 14 +- .../detail/allocator/tagged_allocator.h | 11 +- .../detail/allocator/tagged_allocator.inl | 79 +- .../detail/allocator/temporary_allocator.h | 7 - .../detail/allocator/temporary_allocator.inl | 66 +- .../detail/allocator/value_initialize_range.h | 8 - .../allocator/value_initialize_range.inl | 6 +- .../detail/allocator_aware_execution_policy.h | 4 - projects/rocthrust/thrust/detail/attributes.h | 40 + .../rocthrust/thrust/detail/binary_search.inl | 581 ++-- .../thrust/detail/caching_allocator.h | 7 - .../thrust/detail/complex/arithmetic.h | 84 +- .../rocthrust/thrust/detail/complex/c99math.h | 64 +- .../rocthrust/thrust/detail/complex/ccosh.h | 2 - .../rocthrust/thrust/detail/complex/ccoshf.h | 2 - .../rocthrust/thrust/detail/complex/cexp.h | 4 +- .../rocthrust/thrust/detail/complex/cexpf.h | 12 +- .../rocthrust/thrust/detail/complex/clog.h | 14 +- .../rocthrust/thrust/detail/complex/clogf.h | 14 +- .../thrust/detail/complex/complex.inl | 60 +- .../rocthrust/thrust/detail/complex/cpow.h | 21 +- .../rocthrust/thrust/detail/complex/csinh.h | 1 - .../rocthrust/thrust/detail/complex/csinhf.h | 2 - projects/rocthrust/thrust/detail/config.h | 12 +- .../rocthrust/thrust/detail/config/compiler.h | 116 +- .../thrust/detail/config/compiler_fence.h | 71 + .../rocthrust/thrust/detail/config/config.h | 14 +- .../thrust/detail/config/cpp_compatibility.h | 123 + .../thrust/detail/config/cpp_dialect.h | 149 +- .../thrust/detail/config/deprecated.h | 70 + .../thrust/detail/config/device_system.h | 14 +- .../thrust/detail/config/diagnostic.h | 99 + .../thrust/detail/config/execution_space.h | 51 + .../thrust/detail/config/global_workarounds.h | 26 + .../thrust/detail/config/host_system.h | 14 +- .../rocthrust/thrust/detail/config/libcxx.h | 70 +- .../thrust/detail/config/memory_resource.h | 3 + .../thrust/detail/config/namespace.h | 114 +- .../thrust/detail/config/preprocessor.h | 20 + .../std/__cccl => config}/rtti.h | 46 +- .../thrust/detail/config/simple_defines.h | 19 +- .../thrust/detail/config/visibility.h | 22 + .../thrust/detail/contiguous_storage.h | 126 +- .../thrust/detail/contiguous_storage.inl | 188 +- projects/rocthrust/thrust/detail/copy.h | 7 - projects/rocthrust/thrust/detail/copy.inl | 106 +- projects/rocthrust/thrust/detail/copy_if.h | 7 - projects/rocthrust/thrust/detail/copy_if.inl | 79 +- projects/rocthrust/thrust/detail/count.h | 16 +- projects/rocthrust/thrust/detail/count.inl | 40 +- .../thrust/detail/cpp_version_check.h | 30 + .../dependencies_aware_execution_policy.h | 85 + .../rocthrust/thrust/detail/device_delete.inl | 20 +- .../rocthrust/thrust/detail/device_free.inl | 10 +- .../rocthrust/thrust/detail/device_malloc.inl | 17 +- .../rocthrust/thrust/detail/device_new.inl | 25 +- projects/rocthrust/thrust/detail/distance.inl | 19 +- projects/rocthrust/thrust/detail/equal.inl | 47 +- .../rocthrust/thrust/detail/event_error.h | 18 +- .../thrust/detail/execute_with_allocator.h | 42 + .../detail/execute_with_allocator_fwd.h | 52 +- .../thrust/detail/execute_with_dependencies.h | 226 ++ projects/rocthrust/thrust/detail/extrema.inl | 79 +- projects/rocthrust/thrust/detail/fill.inl | 50 +- projects/rocthrust/thrust/detail/find.inl | 63 +- projects/rocthrust/thrust/detail/for_each.inl | 52 +- .../rocthrust/thrust/detail/functional.inl | 138 + .../thrust/detail/functional/actor.h | 4 +- .../detail/functional/address_stability.h | 133 + projects/rocthrust/thrust/detail/gather.inl | 158 +- projects/rocthrust/thrust/detail/generate.inl | 56 +- .../thrust/detail/get_iterator_value.h | 13 +- .../rocthrust/thrust/detail/inner_product.inl | 97 +- .../rocthrust/thrust/detail/integer_math.h | 36 +- .../rocthrust/thrust/detail/integer_traits.h | 120 + .../thrust/detail/internal_functional.h | 76 +- .../detail/libcxx_wrapper/__cccl_config.h | 38 - .../__functional/address_stability.h | 113 - .../libcxx_wrapper/__functional/maximum.h | 64 - .../libcxx_wrapper/__functional/minimum.h | 64 - .../libcxx_wrapper/std/__cccl/attributes.h | 81 - .../libcxx_wrapper/std/__cccl/builtin.h | 29 - .../libcxx_wrapper/std/__cccl/compiler.h | 217 -- .../libcxx_wrapper/std/__cccl/cuda_toolkit.h | 54 - .../libcxx_wrapper/std/__cccl/deprecated.h | 75 - .../libcxx_wrapper/std/__cccl/diagnostic.h | 134 - .../libcxx_wrapper/std/__cccl/dialect.h | 87 - .../libcxx_wrapper/std/__cccl/exceptions.h | 42 - .../std/__cccl/execution_space.h | 60 - .../libcxx_wrapper/std/__cccl/preprocessor.h | 736 ----- .../std/__cccl/sequence_access.h | 98 - .../libcxx_wrapper/std/__cccl/unreachable.h | 42 - .../libcxx_wrapper/std/__cccl/visibility.h | 37 - .../std/__exception/terminate.h | 67 - .../std/__functional/identity.h | 51 - .../libcxx_wrapper/std/__functional/invoke.h | 176 - .../libcxx_wrapper/std/__functional/not_fn.h | 145 - .../libcxx_wrapper/std/__iterator/concepts.h | 61 - .../std/__iterator/iterator_traits.h | 90 - .../std/__memory/construct_at.h | 153 - .../std/__type_traits/conditional.h | 58 - .../std/__type_traits/conjunction.h | 54 - .../std/__type_traits/is_reference_wrapper.h | 41 - .../libcxx_wrapper/std/__type_traits/nat.h | 36 - .../std/__type_traits/type_identity.h | 41 - .../std/detail/libcxx/include/__config.h | 36 - .../std/detail/libcxx/include/stdexcept.h | 56 - projects/rocthrust/thrust/detail/logical.inl | 50 +- .../rocthrust/thrust/detail/malloc_and_free.h | 6 +- .../thrust/detail/memory_algorithms.h | 60 +- .../rocthrust/thrust/detail/memory_wrapper.h | 10 - projects/rocthrust/thrust/detail/merge.inl | 267 +- projects/rocthrust/thrust/detail/minmax.h | 47 + projects/rocthrust/thrust/detail/mismatch.inl | 61 +- .../thrust/detail/modern_gcc_required.h | 25 + projects/rocthrust/thrust/detail/mpl/math.h | 8 - .../rocthrust/thrust/detail/numeric_traits.h | 114 + .../rocthrust/thrust/detail/numeric_wrapper.h | 10 - projects/rocthrust/thrust/detail/nv_target.h | 32 + .../rocthrust/thrust/detail/overlapped_copy.h | 9 +- projects/rocthrust/thrust/detail/pair.inl | 2 - .../rocthrust/thrust/detail/partition.inl | 371 ++- projects/rocthrust/thrust/detail/pointer.h | 26 +- projects/rocthrust/thrust/detail/pointer.inl | 5 +- .../rocthrust/thrust/detail/preprocessor.h | 543 +++- .../thrust/detail/range/head_flags.h | 7 +- .../thrust/detail/range/tail_flags.h | 12 +- .../thrust/detail/raw_pointer_cast.h | 7 - .../thrust/detail/raw_reference_cast.h | 135 +- projects/rocthrust/thrust/detail/reduce.inl | 289 +- projects/rocthrust/thrust/detail/reference.h | 6 +- .../detail/reference_forward_declaration.h | 7 - projects/rocthrust/thrust/detail/remove.inl | 194 +- projects/rocthrust/thrust/detail/replace.inl | 190 +- projects/rocthrust/thrust/detail/reverse.inl | 54 +- projects/rocthrust/thrust/detail/scan.inl | 513 +-- projects/rocthrust/thrust/detail/scatter.inl | 141 +- .../rocthrust/thrust/detail/select_system.h | 6 +- projects/rocthrust/thrust/detail/seq.h | 11 +- projects/rocthrust/thrust/detail/sequence.inl | 68 +- .../thrust/detail/set_operations.inl | 1097 +++---- projects/rocthrust/thrust/detail/shuffle.inl | 43 +- projects/rocthrust/thrust/detail/sort.inl | 300 +- .../rocthrust/thrust/detail/static_assert.h | 8 +- projects/rocthrust/thrust/detail/static_map.h | 8 - .../rocthrust/thrust/detail/swap_ranges.inl | 38 +- projects/rocthrust/thrust/detail/tabulate.inl | 33 +- .../rocthrust/thrust/detail/temporary_array.h | 2 +- .../thrust/detail/temporary_buffer.h | 7 - .../rocthrust/thrust/detail/transform.inl | 250 +- .../thrust/detail/transform_reduce.inl | 52 +- .../thrust/detail/trivial_sequence.h | 13 +- projects/rocthrust/thrust/detail/tuple.inl | 56 +- .../thrust/detail/tuple_algorithms.h | 68 + .../thrust/detail/tuple_meta_transform.h | 8 - .../rocthrust/thrust/detail/tuple_transform.h | 65 + .../rocthrust/thrust/detail/type_deduction.h | 45 +- .../rocthrust/thrust/detail/type_traits.h | 125 +- .../type_traits/is_metafunction_defined.h | 8 - .../detail/type_traits/is_thrust_pointer.h | 65 - .../iterator/is_discard_iterator.h | 7 - .../type_traits/iterator/is_output_iterator.h | 34 +- .../detail/type_traits/pointer_traits.h | 33 +- .../thrust/detail/uninitialized_copy.inl | 57 +- .../thrust/detail/uninitialized_fill.inl | 53 +- projects/rocthrust/thrust/detail/unique.inl | 345 +- .../rocthrust/thrust/detail/use_default.h | 8 - projects/rocthrust/thrust/detail/util/align.h | 8 - .../rocthrust/thrust/detail/vector_base.h | 25 +- .../rocthrust/thrust/detail/vector_base.inl | 206 +- projects/rocthrust/thrust/device_allocator.h | 9 +- projects/rocthrust/thrust/device_delete.h | 7 - projects/rocthrust/thrust/device_free.h | 7 - projects/rocthrust/thrust/device_malloc.h | 15 +- projects/rocthrust/thrust/device_new.h | 8 - projects/rocthrust/thrust/device_ptr.h | 3 +- projects/rocthrust/thrust/distance.h | 9 +- projects/rocthrust/thrust/equal.h | 19 +- projects/rocthrust/thrust/event.h | 25 + projects/rocthrust/thrust/execution_policy.h | 8 +- projects/rocthrust/thrust/extrema.h | 211 +- projects/rocthrust/thrust/fill.h | 7 - projects/rocthrust/thrust/find.h | 23 +- projects/rocthrust/thrust/for_each.h | 46 +- projects/rocthrust/thrust/functional.h | 1585 ++++++++- projects/rocthrust/thrust/future.h | 195 ++ projects/rocthrust/thrust/gather.h | 15 +- projects/rocthrust/thrust/generate.h | 27 +- projects/rocthrust/thrust/inner_product.h | 31 +- .../thrust/iterator/constant_iterator.h | 277 +- .../thrust/iterator/counting_iterator.h | 318 +- .../thrust/iterator/detail/any_assign.h | 23 +- .../thrust/iterator/detail/any_system_tag.h | 2 +- .../iterator/detail/constant_iterator_base.h | 62 + .../iterator/detail/counting_iterator.inl | 129 + .../iterator/detail/device_system_tag.h | 8 - .../iterator/detail/discard_iterator_base.h | 65 + .../iterator/detail/distance_from_result.h | 50 + .../thrust/iterator/detail/host_system_tag.h | 8 - .../iterator/detail/is_iterator_category.h | 58 + .../iterator/detail/iterator_adaptor_base.h | 47 +- .../detail/iterator_category_to_system.h | 48 +- .../detail/iterator_category_to_traversal.h | 83 +- ...rator_category_with_system_and_traversal.h | 15 +- .../detail/iterator_facade_category.h | 178 +- .../thrust/iterator/detail/iterator_traits.h | 95 + .../iterator/detail/iterator_traits.inl | 129 + .../iterator/detail/iterator_traversal_tags.h | 8 - .../thrust/iterator/detail/join_iterator.h | 126 + .../thrust/iterator/detail/minimum_category.h | 49 + .../thrust/iterator/detail/minimum_system.h | 52 +- .../thrust/iterator/detail/normal_iterator.h | 20 +- .../detail/permutation_iterator_base.h | 51 + .../rocthrust/thrust/iterator/detail/retag.h | 88 +- .../iterator/detail/reverse_iterator.inl | 90 + .../iterator/detail/reverse_iterator_base.h | 40 + .../detail/tabulate_output_iterator.inl | 61 + .../thrust/iterator/detail/tagged_iterator.h | 53 +- .../transform_input_output_iterator.inl | 111 + .../iterator/detail/transform_iterator.inl | 100 + .../detail/transform_output_iterator.inl | 85 + .../detail/tuple_of_iterator_references.h | 44 +- .../iterator/detail/universal_categories.h | 125 + .../thrust/iterator/detail/zip_iterator.inl | 128 + .../iterator/detail/zip_iterator_base.h | 287 ++ .../thrust/iterator/discard_iterator.h | 210 +- .../thrust/iterator/iterator_adaptor.h | 5 +- .../thrust/iterator/iterator_categories.h | 2 + .../thrust/iterator/iterator_facade.h | 30 +- .../thrust/iterator/iterator_traits.h | 176 +- .../thrust/iterator/offset_iterator.h | 201 -- .../thrust/iterator/permutation_iterator.h | 248 +- projects/rocthrust/thrust/iterator/retag.h | 7 - .../thrust/iterator/reverse_iterator.h | 316 +- .../iterator/tabulate_output_iterator.h | 177 +- .../transform_input_output_iterator.h | 251 +- .../thrust/iterator/transform_iterator.h | 480 +-- .../iterator/transform_output_iterator.h | 221 +- .../rocthrust/thrust/iterator/zip_iterator.h | 756 +---- projects/rocthrust/thrust/limits.h | 30 + projects/rocthrust/thrust/logical.h | 4 +- projects/rocthrust/thrust/merge.h | 16 +- projects/rocthrust/thrust/mismatch.h | 7 - projects/rocthrust/thrust/mr/allocator.h | 14 +- .../thrust/mr/device_memory_resource.h | 8 - projects/rocthrust/thrust/mr/disjoint_pool.h | 15 +- .../rocthrust/thrust/mr/disjoint_sync_pool.h | 13 +- .../rocthrust/thrust/mr/disjoint_tls_pool.h | 7 - .../thrust/mr/fancy_pointer_resource.h | 7 - .../thrust/mr/host_memory_resource.h | 8 - .../rocthrust/thrust/mr/memory_resource.h | 9 +- projects/rocthrust/thrust/mr/new.h | 6 +- .../rocthrust/thrust/mr/polymorphic_adaptor.h | 8 - projects/rocthrust/thrust/mr/pool.h | 12 +- projects/rocthrust/thrust/mr/sync_pool.h | 13 +- projects/rocthrust/thrust/mr/tls_pool.h | 7 - .../thrust/mr/universal_memory_resource.h | 8 - projects/rocthrust/thrust/mr/validator.h | 12 +- projects/rocthrust/thrust/optional.h | 2879 +++++++++++++++++ projects/rocthrust/thrust/pair.h | 6 +- projects/rocthrust/thrust/partition.h | 40 +- .../rocthrust/thrust/per_device_resource.h | 9 +- projects/rocthrust/thrust/random.h | 11 +- .../random/detail/discard_block_engine.inl | 154 +- .../detail/linear_congruential_engine.inl | 122 +- .../detail/linear_feedback_shift_engine.inl | 114 +- .../linear_feedback_shift_engine_wordmask.h | 8 - projects/rocthrust/thrust/random/detail/mod.h | 12 +- .../random/detail/normal_distribution.inl | 1 + .../random/detail/normal_distribution_base.h | 8 +- .../thrust/random/detail/random_core_access.h | 8 - .../detail/subtract_with_carry_engine.inl | 149 +- .../detail/uniform_int_distribution.inl | 186 +- .../detail/uniform_real_distribution.inl | 187 +- .../random/detail/xor_combine_engine.inl | 178 +- .../random/detail/xor_combine_engine_max.h | 8 - .../thrust/random/normal_distribution.h | 21 +- .../thrust/random/uniform_int_distribution.h | 26 +- .../thrust/random/uniform_real_distribution.h | 21 +- projects/rocthrust/thrust/reduce.h | 32 +- projects/rocthrust/thrust/remove.h | 16 +- projects/rocthrust/thrust/replace.h | 39 +- projects/rocthrust/thrust/reverse.h | 7 - projects/rocthrust/thrust/scan.h | 60 +- projects/rocthrust/thrust/scatter.h | 17 +- projects/rocthrust/thrust/sequence.h | 7 - projects/rocthrust/thrust/sequence_access.h | 82 + projects/rocthrust/thrust/set_operations.h | 16 +- projects/rocthrust/thrust/shuffle.h | 8 - projects/rocthrust/thrust/sort.h | 29 +- .../system/cpp/detail/adjacent_difference.h | 8 - .../thrust/system/cpp/detail/assign_value.h | 8 - .../thrust/system/cpp/detail/binary_search.h | 10 - .../rocthrust/thrust/system/cpp/detail/copy.h | 8 - .../thrust/system/cpp/detail/copy_if.h | 8 - .../thrust/system/cpp/detail/count.h | 8 - .../thrust/system/cpp/detail/equal.h | 8 - .../thrust/system/cpp/detail/extrema.h | 8 - .../rocthrust/thrust/system/cpp/detail/fill.h | 8 - .../rocthrust/thrust/system/cpp/detail/find.h | 8 - .../thrust/system/cpp/detail/for_each.h | 8 - .../thrust/system/cpp/detail/gather.h | 8 - .../thrust/system/cpp/detail/generate.h | 8 - .../thrust/system/cpp/detail/get_value.h | 8 - .../thrust/system/cpp/detail/inner_product.h | 8 - .../thrust/system/cpp/detail/iter_swap.h | 8 - .../thrust/system/cpp/detail/logical.h | 8 - .../system/cpp/detail/malloc_and_free.h | 8 - .../thrust/system/cpp/detail/memory.inl | 17 +- .../thrust/system/cpp/detail/merge.h | 8 - .../thrust/system/cpp/detail/mismatch.h | 8 - .../rocthrust/thrust/system/cpp/detail/par.h | 9 +- .../thrust/system/cpp/detail/partition.h | 8 - .../system/cpp/detail/per_device_resource.h | 8 - .../thrust/system/cpp/detail/reduce.h | 8 - .../thrust/system/cpp/detail/reduce_by_key.h | 8 - .../thrust/system/cpp/detail/remove.h | 8 - .../thrust/system/cpp/detail/replace.h | 8 - .../thrust/system/cpp/detail/reverse.h | 8 - .../rocthrust/thrust/system/cpp/detail/scan.h | 8 - .../thrust/system/cpp/detail/scan_by_key.h | 8 - .../thrust/system/cpp/detail/scatter.h | 8 - .../thrust/system/cpp/detail/sequence.h | 8 - .../thrust/system/cpp/detail/set_operations.h | 8 - .../rocthrust/thrust/system/cpp/detail/sort.h | 8 - .../thrust/system/cpp/detail/swap_ranges.h | 8 - .../thrust/system/cpp/detail/tabulate.h | 8 - .../system/cpp/detail/temporary_buffer.h | 8 - .../thrust/system/cpp/detail/transform.h | 8 - .../system/cpp/detail/transform_reduce.h | 8 - .../thrust/system/cpp/detail/transform_scan.h | 8 - .../system/cpp/detail/uninitialized_copy.h | 8 - .../system/cpp/detail/uninitialized_fill.h | 8 - .../thrust/system/cpp/detail/unique.h | 8 - .../thrust/system/cpp/detail/unique_by_key.h | 8 - .../thrust/system/cpp/detail/vector.inl | 155 +- .../thrust/system/cpp/execution_policy.h | 10 +- .../thrust/system/cpp/memory_resource.h | 4 +- projects/rocthrust/thrust/system/cpp/vector.h | 5 +- .../rocthrust/thrust/system/cuda/config.h | 49 +- .../system/cuda/detail/adjacent_difference.h | 28 +- .../thrust/system/cuda/detail/assign_value.h | 65 +- .../thrust/system/cuda/detail/async/copy.h | 362 +++ .../system/cuda/detail/async/customization.h | 113 + .../system/cuda/detail/async/exclusive_scan.h | 166 + .../system/cuda/detail/async/for_each.h | 136 + .../system/cuda/detail/async/inclusive_scan.h | 255 ++ .../thrust/system/cuda/detail/async/reduce.h | 236 ++ .../{core/load_iterator.h => async/scan.h} | 35 +- .../thrust/system/cuda/detail/async/sort.h | 328 ++ .../system/cuda/detail/async/transform.h | 141 + .../thrust/system/cuda/detail/binary_search.h | 10 - .../thrust/system/cuda/detail/cdp_dispatch.h | 57 +- .../thrust/system/cuda/detail/copy.h | 2 +- .../thrust/system/cuda/detail/copy_if.h | 103 +- .../system/cuda/detail/core/agent_launcher.h | 51 +- .../cuda/detail/core/triple_chevron_launch.h | 31 +- .../thrust/system/cuda/detail/core/util.h | 281 +- .../thrust/system/cuda/detail/count.h | 8 +- .../thrust/system/cuda/detail/cross_system.h | 1 - .../thrust/system/cuda/detail/dispatch.h | 47 +- .../thrust/system/cuda/detail/equal.h | 4 +- .../thrust/system/cuda/detail/error.inl | 73 +- .../system/cuda/detail/execution_policy.h | 11 +- .../thrust/system/cuda/detail/extrema.h | 51 +- .../thrust/system/cuda/detail/fill.h | 2 +- .../thrust/system/cuda/detail/find.h | 13 +- .../thrust/system/cuda/detail/for_each.h | 4 +- .../thrust/system/cuda/detail/future.inl | 1235 +++++++ .../thrust/system/cuda/detail/gather.h | 2 +- .../thrust/system/cuda/detail/generate.h | 2 +- .../thrust/system/cuda/detail/get_value.h | 52 +- .../thrust/system/cuda/detail/inner_product.h | 3 +- .../cuda/detail/internal/copy_cross_system.h | 8 +- .../detail/internal/copy_device_to_device.h | 4 +- .../thrust/system/cuda/detail/iter_swap.h | 4 +- .../thrust/system/cuda/detail/logical.h | 8 - .../cuda/detail/make_unsigned_special.h | 8 - .../system/cuda/detail/malloc_and_free.h | 34 +- .../thrust/system/cuda/detail/memory.inl | 24 +- .../thrust/system/cuda/detail/merge.h | 10 +- .../thrust/system/cuda/detail/mismatch.h | 6 +- .../rocthrust/thrust/system/cuda/detail/par.h | 7 + .../thrust/system/cuda/detail/par_to_seq.h | 8 - .../thrust/system/cuda/detail/parallel_for.h | 2 +- .../thrust/system/cuda/detail/partition.h | 60 +- .../system/cuda/detail/per_device_resource.h | 2 +- .../thrust/system/cuda/detail/reduce.h | 88 +- .../thrust/system/cuda/detail/reduce_by_key.h | 140 +- .../thrust/system/cuda/detail/remove.h | 6 +- .../thrust/system/cuda/detail/replace.h | 6 +- .../thrust/system/cuda/detail/reverse.h | 6 +- .../thrust/system/cuda/detail/scan.h | 42 +- .../thrust/system/cuda/detail/scan_by_key.h | 8 +- .../thrust/system/cuda/detail/scatter.h | 2 +- .../thrust/system/cuda/detail/sequence.h | 8 - .../system/cuda/detail/set_operations.h | 176 +- .../thrust/system/cuda/detail/sort.h | 83 +- .../thrust/system/cuda/detail/swap_ranges.h | 8 +- .../thrust/system/cuda/detail/tabulate.h | 4 +- .../system/cuda/detail/temporary_buffer.h | 115 +- .../thrust/system/cuda/detail/transform.h | 23 +- .../system/cuda/detail/transform_reduce.h | 6 +- .../system/cuda/detail/transform_scan.h | 12 +- .../system/cuda/detail/uninitialized_copy.h | 6 +- .../system/cuda/detail/uninitialized_fill.h | 4 +- .../thrust/system/cuda/detail/unique.h | 110 +- .../thrust/system/cuda/detail/unique_by_key.h | 17 +- .../thrust/system/cuda/detail/util.h | 383 ++- projects/rocthrust/thrust/system/cuda/error.h | 9 +- .../thrust/system/cuda/execution_policy.h | 7 - .../rocthrust/thrust/system/cuda/future.h | 79 + .../thrust/system/cuda/memory_resource.h | 9 - .../system/detail/adl/adjacent_difference.h | 10 +- .../thrust/system/detail/adl/assign_value.h | 10 +- .../thrust/system/detail/adl/async/copy.h | 33 + .../thrust/system/detail/adl/async/for_each.h | 33 + .../thrust/system/detail/adl/async/reduce.h | 33 + .../thrust/system/detail/adl/async/scan.h | 33 + .../thrust/system/detail/adl/async/sort.h | 33 + .../system/detail/adl/async/transform.h | 33 + .../thrust/system/detail/adl/binary_search.h | 10 +- .../rocthrust/thrust/system/detail/adl/copy.h | 10 +- .../thrust/system/detail/adl/copy_if.h | 11 +- .../thrust/system/detail/adl/count.h | 10 +- .../thrust/system/detail/adl/equal.h | 10 +- .../thrust/system/detail/adl/extrema.h | 10 +- .../rocthrust/thrust/system/detail/adl/fill.h | 10 +- .../rocthrust/thrust/system/detail/adl/find.h | 10 +- .../thrust/system/detail/adl/for_each.h | 10 +- .../thrust/system/detail/adl/gather.h | 10 +- .../thrust/system/detail/adl/generate.h | 10 +- .../thrust/system/detail/adl/get_value.h | 10 +- .../thrust/system/detail/adl/inner_product.h | 10 +- .../thrust/system/detail/adl/iter_swap.h | 10 +- .../thrust/system/detail/adl/logical.h | 10 +- .../system/detail/adl/malloc_and_free.h | 10 +- .../thrust/system/detail/adl/merge.h | 10 +- .../thrust/system/detail/adl/mismatch.h | 10 +- .../thrust/system/detail/adl/partition.h | 10 +- .../system/detail/adl/per_device_resource.h | 9 - .../thrust/system/detail/adl/reduce.h | 10 +- .../thrust/system/detail/adl/reduce_by_key.h | 10 +- .../thrust/system/detail/adl/remove.h | 10 +- .../thrust/system/detail/adl/replace.h | 10 +- .../thrust/system/detail/adl/reverse.h | 10 +- .../rocthrust/thrust/system/detail/adl/scan.h | 10 +- .../thrust/system/detail/adl/scan_by_key.h | 10 +- .../thrust/system/detail/adl/scatter.h | 10 +- .../thrust/system/detail/adl/sequence.h | 10 +- .../thrust/system/detail/adl/set_operations.h | 10 +- .../rocthrust/thrust/system/detail/adl/sort.h | 10 +- .../thrust/system/detail/adl/swap_ranges.h | 10 +- .../thrust/system/detail/adl/tabulate.h | 10 +- .../system/detail/adl/temporary_buffer.h | 10 +- .../thrust/system/detail/adl/transform.h | 10 +- .../system/detail/adl/transform_reduce.h | 10 +- .../thrust/system/detail/adl/transform_scan.h | 10 +- .../system/detail/adl/uninitialized_copy.h | 10 +- .../system/detail/adl/uninitialized_fill.h | 10 +- .../thrust/system/detail/adl/unique.h | 10 +- .../thrust/system/detail/adl/unique_by_key.h | 10 +- .../thrust/system/detail/bad_alloc.h | 12 +- .../rocthrust/thrust/system/detail/errno.h | 8 - .../thrust/system/detail/error_code.inl | 8 +- .../thrust/system/detail/error_condition.inl | 8 +- .../detail/generic/adjacent_difference.h | 7 - .../detail/generic/adjacent_difference.inl | 42 +- .../thrust/system/detail/generic/advance.h | 8 - .../thrust/system/detail/generic/advance.inl | 69 +- .../system/detail/generic/binary_search.h | 7 - .../system/detail/generic/binary_search.inl | 382 +-- .../thrust/system/detail/generic/copy.h | 7 - .../thrust/system/detail/generic/copy.inl | 1 - .../thrust/system/detail/generic/copy_if.h | 7 - .../thrust/system/detail/generic/copy_if.inl | 12 +- .../thrust/system/detail/generic/count.h | 11 +- .../thrust/system/detail/generic/count.inl | 51 +- .../thrust/system/detail/generic/distance.h | 9 +- .../thrust/system/detail/generic/distance.inl | 44 +- .../thrust/system/detail/generic/equal.h | 7 - .../thrust/system/detail/generic/extrema.h | 7 - .../thrust/system/detail/generic/extrema.inl | 18 +- .../thrust/system/detail/generic/fill.h | 8 - .../thrust/system/detail/generic/find.h | 7 - .../thrust/system/detail/generic/find.inl | 92 +- .../thrust/system/detail/generic/for_each.h | 15 +- .../thrust/system/detail/generic/gather.h | 7 - .../thrust/system/detail/generic/gather.inl | 1 - .../thrust/system/detail/generic/generate.h | 7 - .../thrust/system/detail/generic/generate.inl | 73 +- .../system/detail/generic/inner_product.h | 7 - .../system/detail/generic/inner_product.inl | 55 +- .../thrust/system/detail/generic/logical.h | 7 - .../thrust/system/detail/generic/memory.h | 7 - .../thrust/system/detail/generic/memory.inl | 86 +- .../thrust/system/detail/generic/merge.h | 7 - .../thrust/system/detail/generic/merge.inl | 147 +- .../thrust/system/detail/generic/mismatch.h | 7 - .../thrust/system/detail/generic/mismatch.inl | 54 +- .../thrust/system/detail/generic/partition.h | 7 - .../system/detail/generic/partition.inl | 239 +- .../detail/generic/per_device_resource.h | 7 - .../thrust/system/detail/generic/reduce.h | 9 +- .../thrust/system/detail/generic/reduce.inl | 52 +- .../system/detail/generic/reduce_by_key.h | 7 - .../system/detail/generic/reduce_by_key.inl | 204 +- .../thrust/system/detail/generic/remove.h | 7 - .../thrust/system/detail/generic/remove.inl | 124 +- .../thrust/system/detail/generic/replace.h | 7 - .../thrust/system/detail/generic/replace.inl | 163 +- .../thrust/system/detail/generic/reverse.h | 7 - .../thrust/system/detail/generic/reverse.inl | 46 +- .../detail/generic/scalar/binary_search.h | 7 - .../detail/generic/scalar/binary_search.inl | 100 +- .../thrust/system/detail/generic/scan.h | 7 - .../thrust/system/detail/generic/scan.inl | 101 +- .../system/detail/generic/scan_by_key.h | 7 - .../system/detail/generic/scan_by_key.inl | 241 +- .../thrust/system/detail/generic/scatter.h | 7 - .../thrust/system/detail/generic/scatter.inl | 1 - .../system/detail/generic/select_system.h | 7 - .../detail/generic/select_system_exists.h | 8 - .../thrust/system/detail/generic/sequence.h | 7 - .../thrust/system/detail/generic/sequence.inl | 57 +- .../system/detail/generic/set_operations.h | 7 - .../system/detail/generic/set_operations.inl | 601 ++-- .../thrust/system/detail/generic/shuffle.h | 8 - .../thrust/system/detail/generic/shuffle.inl | 151 +- .../thrust/system/detail/generic/sort.h | 7 - .../thrust/system/detail/generic/sort.inl | 217 +- .../system/detail/generic/swap_ranges.h | 7 - .../thrust/system/detail/generic/tabulate.h | 7 - .../thrust/system/detail/generic/tabulate.inl | 29 +- .../thrust/system/detail/generic/tag.h | 8 - .../system/detail/generic/temporary_buffer.h | 7 - .../detail/generic/temporary_buffer.inl | 43 +- .../thrust/system/detail/generic/transform.h | 7 - .../system/detail/generic/transform.inl | 203 +- .../system/detail/generic/transform_reduce.h | 7 - .../detail/generic/transform_reduce.inl | 44 +- .../system/detail/generic/transform_scan.inl | 9 +- .../detail/generic/uninitialized_copy.h | 7 - .../detail/generic/uninitialized_copy.inl | 14 +- .../detail/generic/uninitialized_fill.h | 7 - .../detail/generic/uninitialized_fill.inl | 8 +- .../thrust/system/detail/generic/unique.h | 11 +- .../thrust/system/detail/generic/unique.inl | 117 +- .../system/detail/generic/unique_by_key.h | 7 - .../system/detail/generic/unique_by_key.inl | 143 +- .../thrust/system/detail/internal/decompose.h | 16 +- .../detail/sequential/adjacent_difference.h | 9 +- .../system/detail/sequential/assign_value.h | 7 - .../system/detail/sequential/binary_search.h | 12 +- .../thrust/system/detail/sequential/copy.h | 7 - .../thrust/system/detail/sequential/copy.inl | 120 +- .../system/detail/sequential/copy_backward.h | 8 - .../thrust/system/detail/sequential/copy_if.h | 7 - .../thrust/system/detail/sequential/count.h | 8 - .../thrust/system/detail/sequential/equal.h | 8 - .../detail/sequential/execution_policy.h | 4 +- .../thrust/system/detail/sequential/extrema.h | 7 - .../thrust/system/detail/sequential/fill.h | 8 - .../thrust/system/detail/sequential/find.h | 7 - .../system/detail/sequential/for_each.h | 7 - .../thrust/system/detail/sequential/gather.h | 8 - .../system/detail/sequential/general_copy.h | 32 +- .../system/detail/sequential/generate.h | 8 - .../system/detail/sequential/get_value.h | 9 +- .../system/detail/sequential/inner_product.h | 8 - .../system/detail/sequential/insertion_sort.h | 14 +- .../system/detail/sequential/iter_swap.h | 2 +- .../thrust/system/detail/sequential/logical.h | 8 - .../detail/sequential/malloc_and_free.h | 7 - .../thrust/system/detail/sequential/merge.h | 7 - .../thrust/system/detail/sequential/merge.inl | 102 +- .../system/detail/sequential/mismatch.h | 8 - .../system/detail/sequential/partition.h | 6 +- .../detail/sequential/per_device_resource.h | 8 - .../thrust/system/detail/sequential/reduce.h | 7 - .../system/detail/sequential/reduce_by_key.h | 13 +- .../thrust/system/detail/sequential/remove.h | 7 - .../thrust/system/detail/sequential/replace.h | 8 - .../thrust/system/detail/sequential/reverse.h | 8 - .../thrust/system/detail/sequential/scan.h | 12 +- .../system/detail/sequential/scan_by_key.h | 13 +- .../thrust/system/detail/sequential/scatter.h | 8 - .../system/detail/sequential/sequence.h | 8 - .../system/detail/sequential/set_operations.h | 7 - .../thrust/system/detail/sequential/sort.h | 7 - .../thrust/system/detail/sequential/sort.inl | 27 +- .../detail/sequential/stable_merge_sort.h | 7 - .../detail/sequential/stable_merge_sort.inl | 484 +-- .../detail/sequential/stable_primitive_sort.h | 7 - .../sequential/stable_primitive_sort.inl | 4 +- .../detail/sequential/stable_radix_sort.h | 7 - .../detail/sequential/stable_radix_sort.inl | 16 +- .../system/detail/sequential/swap_ranges.h | 8 - .../system/detail/sequential/tabulate.h | 8 - .../detail/sequential/temporary_buffer.h | 8 - .../system/detail/sequential/transform.h | 8 - .../detail/sequential/transform_reduce.h | 8 - .../system/detail/sequential/transform_scan.h | 8 - .../system/detail/sequential/trivial_copy.h | 19 +- .../detail/sequential/uninitialized_copy.h | 8 - .../detail/sequential/uninitialized_fill.h | 8 - .../thrust/system/detail/sequential/unique.h | 15 +- .../system/detail/sequential/unique_by_key.h | 11 +- .../thrust/system/detail/system_error.inl | 77 +- projects/rocthrust/thrust/system/error_code.h | 22 +- projects/rocthrust/thrust/system/hip/config.h | 48 +- .../system/hip/detail/adjacent_difference.h | 22 +- .../thrust/system/hip/detail/assign_value.h | 69 +- .../thrust/system/hip/detail/async/copy.h | 357 ++ .../system/hip/detail/async/customization.h | 112 + .../system/hip/detail/async/exclusive_scan.h | 150 + .../thrust/system/hip/detail/async/for_each.h | 132 + .../system/hip/detail/async/inclusive_scan.h | 221 ++ .../thrust/system/hip/detail/async/reduce.h | 250 ++ .../detail/async/scan.h} | 38 +- .../thrust/system/hip/detail/async/sort.h | 324 ++ .../system/hip/detail/async/transform.h | 136 + .../thrust/system/hip/detail/binary_search.h | 95 +- .../rocthrust/thrust/system/hip/detail/copy.h | 12 +- .../thrust/system/hip/detail/copy_if.h | 31 +- .../thrust/system/hip/detail/count.h | 10 +- .../thrust/system/hip/detail/cross_system.h | 57 +- .../thrust/system/hip/detail/dispatch.h | 249 +- .../thrust/system/hip/detail/equal.h | 6 +- .../thrust/system/hip/detail/error.inl | 90 +- .../system/hip/detail/execution_policy.h | 8 +- .../thrust/system/hip/detail/extrema.h | 42 +- .../rocthrust/thrust/system/hip/detail/fill.h | 4 +- .../rocthrust/thrust/system/hip/detail/find.h | 18 +- .../thrust/system/hip/detail/for_each.h | 6 +- .../thrust/system/hip/detail/future.inl | 1242 +++++++ .../thrust/system/hip/detail/gather.h | 6 +- .../system/hip/detail/general/various.h | 3 +- .../thrust/system/hip/detail/generate.h | 4 +- .../thrust/system/hip/detail/get_value.h | 56 +- .../system/hip/detail/guarded_driver_types.h | 46 + .../hip/detail/guarded_hip_runtime_api.h | 38 + .../thrust/system/hip/detail/inner_product.h | 5 +- .../hip/detail/internal/copy_cross_system.h | 10 +- .../detail/internal/copy_device_to_device.h | 13 +- .../thrust/system/hip/detail/iter_swap.h | 15 +- .../thrust/system/hip/detail/logical.h | 10 +- .../system/hip/detail/make_unsigned_special.h | 10 +- .../system/hip/detail/malloc_and_free.h | 86 +- .../thrust/system/hip/detail/memory.inl | 32 +- .../thrust/system/hip/detail/merge.h | 16 +- .../thrust/system/hip/detail/mismatch.h | 10 +- .../hip/detail}/nv/detail/__preprocessor.h | 13 +- .../hip/detail}/nv/detail/__target_macros.h | 20 +- .../thrust/system/hip/detail/nv/target.h | 21 + .../rocthrust/thrust/system/hip/detail/par.h | 32 +- .../thrust/system/hip/detail/par_to_seq.h | 41 +- .../thrust/system/hip/detail/parallel_for.h | 14 +- .../thrust/system/hip/detail/partition.h | 52 +- .../system/hip/detail/per_device_resource.h | 4 +- .../thrust/system/hip/detail/reduce.h | 18 +- .../thrust/system/hip/detail/reduce_by_key.h | 26 +- .../thrust/system/hip/detail/remove.h | 4 +- .../thrust/system/hip/detail/replace.h | 8 +- .../thrust/system/hip/detail/reverse.h | 10 +- .../rocthrust/thrust/system/hip/detail/scan.h | 26 +- .../thrust/system/hip/detail/scan_by_key.h | 23 +- .../thrust/system/hip/detail/scatter.h | 6 +- .../thrust/system/hip/detail/sequence.h | 10 +- .../thrust/system/hip/detail/set_operations.h | 101 +- .../rocthrust/thrust/system/hip/detail/sort.h | 42 +- .../thrust/system/hip/detail/swap_ranges.h | 18 +- .../thrust/system/hip/detail/tabulate.h | 6 +- .../system/hip/detail/temporary_buffer.h | 121 +- .../thrust/system/hip/detail/terminate.h | 18 +- .../thrust/system/hip/detail/transform.h | 11 +- .../system/hip/detail/transform_reduce.h | 8 +- .../thrust/system/hip/detail/transform_scan.h | 25 +- .../system/hip/detail/uninitialized_copy.h | 8 +- .../system/hip/detail/uninitialized_fill.h | 6 +- .../thrust/system/hip/detail/unique.h | 25 +- .../thrust/system/hip/detail/unique_by_key.h | 19 +- .../rocthrust/thrust/system/hip/detail/util.h | 426 ++- projects/rocthrust/thrust/system/hip/error.h | 48 +- .../thrust/system/hip/execution_policy.h | 10 +- projects/rocthrust/thrust/system/hip/future.h | 80 + .../system/hip/hipstdpar/impl/batch.hpp | 19 +- .../thrust/system/hip/hipstdpar/impl/copy.hpp | 11 +- .../system/hip/hipstdpar/impl/generation.hpp | 23 +- .../system/hip/hipstdpar/impl/hipstd.hpp | 14 +- .../impl/lexicographical_comparison.hpp | 7 +- .../system/hip/hipstdpar/impl/merge.hpp | 7 +- .../system/hip/hipstdpar/impl/min_max.hpp | 15 +- .../system/hip/hipstdpar/impl/numeric.hpp | 57 +- .../hip/hipstdpar/impl/order_changing.hpp | 7 +- .../hip/hipstdpar/impl/partitioning.hpp | 11 +- .../system/hip/hipstdpar/impl/removing.hpp | 19 +- .../system/hip/hipstdpar/impl/search.hpp | 55 +- .../thrust/system/hip/hipstdpar/impl/set.hpp | 28 +- .../system/hip/hipstdpar/impl/sorting.hpp | 31 +- .../thrust/system/hip/hipstdpar/impl/swap.hpp | 5 +- .../hip/hipstdpar/impl/transformation.hpp | 40 +- .../hip/hipstdpar/impl/uninitialized.hpp | 28 +- .../thrust/system/hip/memory_resource.h | 21 +- .../system/omp/detail/adjacent_difference.h | 7 - .../thrust/system/omp/detail/assign_value.h | 8 - .../thrust/system/omp/detail/binary_search.h | 7 - .../rocthrust/thrust/system/omp/detail/copy.h | 7 - .../thrust/system/omp/detail/copy.inl | 147 +- .../thrust/system/omp/detail/copy_if.h | 7 - .../thrust/system/omp/detail/copy_if.inl | 41 +- .../thrust/system/omp/detail/count.h | 8 - .../system/omp/detail/default_decomposition.h | 7 - .../omp/detail/default_decomposition.inl | 19 +- .../thrust/system/omp/detail/equal.h | 8 - .../system/omp/detail/execution_policy.h | 7 - .../thrust/system/omp/detail/extrema.h | 7 - .../rocthrust/thrust/system/omp/detail/fill.h | 8 - .../rocthrust/thrust/system/omp/detail/find.h | 7 - .../thrust/system/omp/detail/for_each.h | 7 - .../thrust/system/omp/detail/for_each.inl | 52 +- .../thrust/system/omp/detail/gather.h | 8 - .../thrust/system/omp/detail/generate.h | 8 - .../thrust/system/omp/detail/get_value.h | 8 - .../thrust/system/omp/detail/inner_product.h | 8 - .../thrust/system/omp/detail/iter_swap.h | 8 - .../thrust/system/omp/detail/logical.h | 8 - .../system/omp/detail/malloc_and_free.h | 8 - .../thrust/system/omp/detail/memory.inl | 26 +- .../thrust/system/omp/detail/merge.h | 8 - .../thrust/system/omp/detail/mismatch.h | 8 - .../rocthrust/thrust/system/omp/detail/par.h | 7 - .../thrust/system/omp/detail/partition.h | 7 - .../thrust/system/omp/detail/partition.inl | 94 +- .../system/omp/detail/per_device_resource.h | 8 - .../thrust/system/omp/detail/reduce.h | 7 - .../thrust/system/omp/detail/reduce.inl | 41 +- .../thrust/system/omp/detail/reduce_by_key.h | 7 - .../system/omp/detail/reduce_by_key.inl | 40 +- .../system/omp/detail/reduce_intervals.h | 7 - .../system/omp/detail/reduce_intervals.inl | 37 +- .../thrust/system/omp/detail/remove.h | 7 - .../thrust/system/omp/detail/remove.inl | 73 +- .../thrust/system/omp/detail/replace.h | 8 - .../thrust/system/omp/detail/reverse.h | 8 - .../rocthrust/thrust/system/omp/detail/scan.h | 8 - .../thrust/system/omp/detail/scan_by_key.h | 8 - .../thrust/system/omp/detail/scatter.h | 8 - .../thrust/system/omp/detail/sequence.h | 8 - .../thrust/system/omp/detail/set_operations.h | 8 - .../rocthrust/thrust/system/omp/detail/sort.h | 7 - .../thrust/system/omp/detail/sort.inl | 221 +- .../thrust/system/omp/detail/swap_ranges.h | 8 - .../thrust/system/omp/detail/tabulate.h | 8 - .../system/omp/detail/temporary_buffer.h | 8 - .../thrust/system/omp/detail/transform.h | 8 - .../system/omp/detail/transform_reduce.h | 8 - .../thrust/system/omp/detail/transform_scan.h | 8 - .../system/omp/detail/uninitialized_copy.h | 8 - .../system/omp/detail/uninitialized_fill.h | 8 - .../thrust/system/omp/detail/unique.h | 9 +- .../thrust/system/omp/detail/unique.inl | 59 +- .../thrust/system/omp/detail/unique_by_key.h | 7 - .../system/omp/detail/unique_by_key.inl | 65 +- .../thrust/system/omp/execution_policy.h | 14 +- .../thrust/system/omp/memory_resource.h | 4 +- projects/rocthrust/thrust/system/omp/vector.h | 5 +- .../rocthrust/thrust/system/system_error.h | 19 +- .../system/tbb/detail/adjacent_difference.h | 7 - .../thrust/system/tbb/detail/assign_value.h | 8 - .../thrust/system/tbb/detail/binary_search.h | 8 - .../rocthrust/thrust/system/tbb/detail/copy.h | 7 - .../thrust/system/tbb/detail/copy.inl | 154 +- .../thrust/system/tbb/detail/copy_if.h | 7 - .../thrust/system/tbb/detail/copy_if.inl | 69 +- .../thrust/system/tbb/detail/count.h | 8 - .../thrust/system/tbb/detail/equal.h | 8 - .../system/tbb/detail/execution_policy.h | 7 - .../thrust/system/tbb/detail/extrema.h | 7 - .../rocthrust/thrust/system/tbb/detail/fill.h | 8 - .../rocthrust/thrust/system/tbb/detail/find.h | 7 - .../thrust/system/tbb/detail/for_each.h | 7 - .../thrust/system/tbb/detail/for_each.inl | 58 +- .../thrust/system/tbb/detail/gather.h | 8 - .../thrust/system/tbb/detail/generate.h | 8 - .../thrust/system/tbb/detail/get_value.h | 8 - .../thrust/system/tbb/detail/inner_product.h | 8 - .../thrust/system/tbb/detail/iter_swap.h | 8 - .../thrust/system/tbb/detail/logical.h | 8 - .../system/tbb/detail/malloc_and_free.h | 8 - .../thrust/system/tbb/detail/memory.inl | 28 +- .../thrust/system/tbb/detail/merge.h | 7 - .../thrust/system/tbb/detail/merge.inl | 198 +- .../thrust/system/tbb/detail/mismatch.h | 8 - .../rocthrust/thrust/system/tbb/detail/par.h | 7 - .../thrust/system/tbb/detail/partition.h | 7 - .../thrust/system/tbb/detail/partition.inl | 92 +- .../system/tbb/detail/per_device_resource.h | 8 - .../thrust/system/tbb/detail/reduce.h | 7 - .../thrust/system/tbb/detail/reduce.inl | 2 +- .../thrust/system/tbb/detail/reduce_by_key.h | 7 - .../system/tbb/detail/reduce_by_key.inl | 44 +- .../system/tbb/detail/reduce_intervals.h | 16 +- .../thrust/system/tbb/detail/remove.h | 7 - .../thrust/system/tbb/detail/remove.inl | 73 +- .../thrust/system/tbb/detail/replace.h | 8 - .../thrust/system/tbb/detail/reverse.h | 8 - .../rocthrust/thrust/system/tbb/detail/scan.h | 7 - .../thrust/system/tbb/detail/scan.inl | 19 +- .../thrust/system/tbb/detail/scan_by_key.h | 8 - .../thrust/system/tbb/detail/scatter.h | 8 - .../thrust/system/tbb/detail/sequence.h | 8 - .../thrust/system/tbb/detail/set_operations.h | 8 - .../rocthrust/thrust/system/tbb/detail/sort.h | 7 - .../thrust/system/tbb/detail/sort.inl | 231 +- .../thrust/system/tbb/detail/swap_ranges.h | 8 - .../thrust/system/tbb/detail/tabulate.h | 8 - .../system/tbb/detail/temporary_buffer.h | 8 - .../thrust/system/tbb/detail/transform.h | 8 - .../system/tbb/detail/transform_reduce.h | 8 - .../thrust/system/tbb/detail/transform_scan.h | 8 - .../system/tbb/detail/uninitialized_copy.h | 8 - .../system/tbb/detail/uninitialized_fill.h | 8 - .../thrust/system/tbb/detail/unique.h | 9 +- .../thrust/system/tbb/detail/unique.inl | 59 +- .../thrust/system/tbb/detail/unique_by_key.h | 7 - .../system/tbb/detail/unique_by_key.inl | 65 +- .../thrust/system/tbb/execution_policy.h | 14 +- .../thrust/system/tbb/memory_resource.h | 4 +- projects/rocthrust/thrust/system/tbb/vector.h | 5 +- projects/rocthrust/thrust/system_error.h | 17 +- projects/rocthrust/thrust/tabulate.h | 17 +- projects/rocthrust/thrust/transform.h | 78 +- projects/rocthrust/thrust/transform_reduce.h | 37 +- projects/rocthrust/thrust/transform_scan.h | 64 +- projects/rocthrust/thrust/tuple.h | 195 +- .../thrust/type_traits/integer_sequence.h | 143 +- .../type_traits/is_contiguous_iterator.h | 129 +- .../thrust/type_traits/is_execution_policy.h | 2 + ...operator_less_or_greater_function_object.h | 17 +- .../is_operator_plus_function_object.h | 11 +- .../type_traits/is_trivially_relocatable.h | 53 +- .../type_traits/logical_metafunctions.h | 26 + .../thrust/type_traits/remove_cvref.h | 43 + .../type_traits/unwrap_contiguous_iterator.h | 94 - .../rocthrust/thrust/type_traits/void_t.h | 63 + .../rocthrust/thrust/uninitialized_copy.h | 15 +- .../rocthrust/thrust/uninitialized_fill.h | 15 +- projects/rocthrust/thrust/unique.h | 41 +- projects/rocthrust/thrust/unique_ptr.h | 222 +- .../rocthrust/thrust/universal_allocator.h | 30 +- projects/rocthrust/thrust/universal_ptr.h | 10 - projects/rocthrust/thrust/universal_vector.h | 21 +- projects/rocthrust/thrust/version.h | 8 +- projects/rocthrust/thrust/zip_function.h | 69 +- 1124 files changed, 44358 insertions(+), 22775 deletions(-) create mode 100644 projects/rocthrust/test/test_async_copy.cpp create mode 100644 projects/rocthrust/test/test_async_for_each.cpp create mode 100644 projects/rocthrust/test/test_async_reduce.cpp create mode 100644 projects/rocthrust/test/test_async_reduce_into.cpp create mode 100644 projects/rocthrust/test/test_async_scan.cpp create mode 100644 projects/rocthrust/test/test_async_sort.cpp create mode 100644 projects/rocthrust/test/test_async_transform.cpp create mode 100644 projects/rocthrust/test/test_dependencies_aware_policies.cpp create mode 100644 projects/rocthrust/test/test_event.cpp create mode 100644 projects/rocthrust/test/test_future.cpp delete mode 100644 projects/rocthrust/test/test_offset_iterator.cpp delete mode 100644 projects/rocthrust/test/test_raw_reference_cast.cpp create mode 100644 projects/rocthrust/test/test_tuple_algorithms.cpp create mode 100644 projects/rocthrust/testing/async/CMakeLists.txt create mode 100644 projects/rocthrust/testing/async/exclusive_scan/CMakeLists.txt create mode 100644 projects/rocthrust/testing/async/exclusive_scan/counting_iterator.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/discard_output.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/large_indices.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/large_types.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/mixed_types.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/mixin.h create mode 100644 projects/rocthrust/testing/async/exclusive_scan/simple.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/stateful_operator.cu create mode 100644 projects/rocthrust/testing/async/exclusive_scan/using_vs_adl.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/CMakeLists.txt create mode 100644 projects/rocthrust/testing/async/inclusive_scan/counting_iterator.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/discard_output.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/large_indices.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/large_types.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/mixed_types.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/mixin.h create mode 100644 projects/rocthrust/testing/async/inclusive_scan/simple.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/stateful_operator.cu create mode 100644 projects/rocthrust/testing/async/inclusive_scan/using_vs_adl.cu create mode 100644 projects/rocthrust/testing/async/mixin.h create mode 100644 projects/rocthrust/testing/async/test_policy_overloads.h create mode 100644 projects/rocthrust/testing/async_copy.cu create mode 100644 projects/rocthrust/testing/async_for_each.cu create mode 100644 projects/rocthrust/testing/async_reduce.cu create mode 100644 projects/rocthrust/testing/async_reduce_into.cu create mode 100644 projects/rocthrust/testing/async_sort.cu create mode 100644 projects/rocthrust/testing/async_transform.cu delete mode 100644 projects/rocthrust/testing/cuda/offset_iterator.cu delete mode 100644 projects/rocthrust/testing/cuda/stream_wrapper.cu delete mode 100644 projects/rocthrust/testing/cuda/transform_iterator.cmake delete mode 100644 projects/rocthrust/testing/cuda/transform_iterator.cu create mode 100644 projects/rocthrust/testing/dependencies_aware_policies.cu create mode 100644 projects/rocthrust/testing/event.cu create mode 100644 projects/rocthrust/testing/future.cu create mode 100644 projects/rocthrust/testing/generate_const_iterators.cu delete mode 100644 projects/rocthrust/testing/offset_iterator.cu delete mode 100644 projects/rocthrust/testing/raw_reference_cast.cu create mode 100644 projects/rocthrust/testing/tuple_algorithms.cu create mode 100644 projects/rocthrust/testing/unittest/runtime_static_assert.h create mode 100644 projects/rocthrust/testing/unittest/util_async.h create mode 100644 projects/rocthrust/testing/unittest_static_assert.cmake create mode 100644 projects/rocthrust/testing/unittest_static_assert.cu create mode 100644 projects/rocthrust/thrust/async/copy.h create mode 100644 projects/rocthrust/thrust/async/for_each.h create mode 100644 projects/rocthrust/thrust/async/reduce.h create mode 100644 projects/rocthrust/thrust/async/scan.h create mode 100644 projects/rocthrust/thrust/async/sort.h create mode 100644 projects/rocthrust/thrust/async/transform.h create mode 100644 projects/rocthrust/thrust/detail/attributes.h create mode 100644 projects/rocthrust/thrust/detail/config/compiler_fence.h create mode 100644 projects/rocthrust/thrust/detail/config/cpp_compatibility.h create mode 100644 projects/rocthrust/thrust/detail/config/deprecated.h create mode 100644 projects/rocthrust/thrust/detail/config/diagnostic.h create mode 100644 projects/rocthrust/thrust/detail/config/execution_space.h create mode 100644 projects/rocthrust/thrust/detail/config/global_workarounds.h create mode 100644 projects/rocthrust/thrust/detail/config/preprocessor.h rename projects/rocthrust/thrust/detail/{libcxx_wrapper/std/__cccl => config}/rtti.h (58%) create mode 100644 projects/rocthrust/thrust/detail/config/visibility.h create mode 100644 projects/rocthrust/thrust/detail/cpp_version_check.h create mode 100644 projects/rocthrust/thrust/detail/dependencies_aware_execution_policy.h create mode 100644 projects/rocthrust/thrust/detail/execute_with_dependencies.h create mode 100644 projects/rocthrust/thrust/detail/functional.inl create mode 100644 projects/rocthrust/thrust/detail/functional/address_stability.h create mode 100644 projects/rocthrust/thrust/detail/integer_traits.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/__cccl_config.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/__functional/address_stability.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/__functional/maximum.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/__functional/minimum.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/attributes.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/builtin.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/compiler.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/cuda_toolkit.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/deprecated.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/diagnostic.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/dialect.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/exceptions.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/execution_space.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/preprocessor.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/sequence_access.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/unreachable.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__cccl/visibility.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__exception/terminate.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__functional/identity.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__functional/invoke.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__functional/not_fn.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__iterator/concepts.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__iterator/iterator_traits.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__memory/construct_at.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__type_traits/conditional.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__type_traits/conjunction.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__type_traits/is_reference_wrapper.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__type_traits/nat.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/__type_traits/type_identity.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/detail/libcxx/include/__config.h delete mode 100644 projects/rocthrust/thrust/detail/libcxx_wrapper/std/detail/libcxx/include/stdexcept.h create mode 100644 projects/rocthrust/thrust/detail/minmax.h create mode 100644 projects/rocthrust/thrust/detail/modern_gcc_required.h create mode 100644 projects/rocthrust/thrust/detail/numeric_traits.h create mode 100644 projects/rocthrust/thrust/detail/nv_target.h create mode 100644 projects/rocthrust/thrust/detail/tuple_algorithms.h create mode 100644 projects/rocthrust/thrust/detail/tuple_transform.h delete mode 100644 projects/rocthrust/thrust/detail/type_traits/is_thrust_pointer.h create mode 100644 projects/rocthrust/thrust/event.h create mode 100644 projects/rocthrust/thrust/future.h create mode 100644 projects/rocthrust/thrust/iterator/detail/constant_iterator_base.h create mode 100644 projects/rocthrust/thrust/iterator/detail/counting_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/discard_iterator_base.h create mode 100644 projects/rocthrust/thrust/iterator/detail/distance_from_result.h create mode 100644 projects/rocthrust/thrust/iterator/detail/is_iterator_category.h create mode 100644 projects/rocthrust/thrust/iterator/detail/iterator_traits.h create mode 100644 projects/rocthrust/thrust/iterator/detail/iterator_traits.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/join_iterator.h create mode 100644 projects/rocthrust/thrust/iterator/detail/minimum_category.h create mode 100644 projects/rocthrust/thrust/iterator/detail/permutation_iterator_base.h create mode 100644 projects/rocthrust/thrust/iterator/detail/reverse_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/reverse_iterator_base.h create mode 100644 projects/rocthrust/thrust/iterator/detail/tabulate_output_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/transform_input_output_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/transform_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/transform_output_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/universal_categories.h create mode 100644 projects/rocthrust/thrust/iterator/detail/zip_iterator.inl create mode 100644 projects/rocthrust/thrust/iterator/detail/zip_iterator_base.h delete mode 100644 projects/rocthrust/thrust/iterator/offset_iterator.h create mode 100644 projects/rocthrust/thrust/limits.h create mode 100644 projects/rocthrust/thrust/optional.h create mode 100644 projects/rocthrust/thrust/sequence_access.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/copy.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/customization.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/exclusive_scan.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/for_each.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/inclusive_scan.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/reduce.h rename projects/rocthrust/thrust/system/cuda/detail/{core/load_iterator.h => async/scan.h} (66%) create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/sort.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/async/transform.h create mode 100644 projects/rocthrust/thrust/system/cuda/detail/future.inl create mode 100644 projects/rocthrust/thrust/system/cuda/future.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/copy.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/for_each.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/reduce.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/scan.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/sort.h create mode 100644 projects/rocthrust/thrust/system/detail/adl/async/transform.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/copy.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/customization.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/exclusive_scan.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/for_each.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/inclusive_scan.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/reduce.h rename projects/rocthrust/thrust/system/{cuda/detail/core/make_load_iterator.h => hip/detail/async/scan.h} (63%) create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/sort.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/async/transform.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/future.inl create mode 100644 projects/rocthrust/thrust/system/hip/detail/guarded_driver_types.h create mode 100644 projects/rocthrust/thrust/system/hip/detail/guarded_hip_runtime_api.h rename projects/rocthrust/thrust/{detail/libcxx_wrapper => system/hip/detail}/nv/detail/__preprocessor.h (89%) rename projects/rocthrust/thrust/{detail/libcxx_wrapper => system/hip/detail}/nv/detail/__target_macros.h (65%) create mode 100644 projects/rocthrust/thrust/system/hip/detail/nv/target.h create mode 100644 projects/rocthrust/thrust/system/hip/future.h create mode 100644 projects/rocthrust/thrust/type_traits/remove_cvref.h delete mode 100644 projects/rocthrust/thrust/type_traits/unwrap_contiguous_iterator.h create mode 100644 projects/rocthrust/thrust/type_traits/void_t.h diff --git a/projects/rocprim/test/rocprim/test_device_segmented_topk.cpp b/projects/rocprim/test/rocprim/test_device_segmented_topk.cpp index 774a3808100c..8ea85172595c 100644 --- a/projects/rocprim/test/rocprim/test_device_segmented_topk.cpp +++ b/projects/rocprim/test/rocprim/test_device_segmented_topk.cpp @@ -910,26 +910,6 @@ void segmented_topk_large_sizes_test(bool debug_synchronous) TEST(RocprimDeviceSegmentedTopkTests, SegmentedTopkLargeSizesStable) { - // Temporarily disable this test on gfx115x on Windows until we can determine the root cause. - // TODO: remove this after the root cause has been found and fixed properly. -#if defined(_WIN32) - rocprim::detail::target_arch arch = rocprim::detail::target_arch::unknown; - if (rocprim::detail::host_target_arch(hipStreamDefault, arch) != HIP_SUCCESS) - { - std::cerr << "Warning: unable to fetch target architecture for disablement check." << std::endl; - } - - const std::set disabled_arches = { - rocprim::detail::target_arch::gfx1150, - rocprim::detail::target_arch::gfx1151, - rocprim::detail::target_arch::gfx1152, - rocprim::detail::target_arch::gfx1153 - }; - - if (disabled_arches.find(arch) != disabled_arches.end()) - GTEST_SKIP() << "Skipping test on gfx1151 Windows systems due to known issues."; -#endif - constexpr bool descending = false; constexpr bool stable = true; using decomposer_t = rocprim::identity_decomposer; @@ -954,26 +934,6 @@ TEST(RocprimDeviceSegmentedTopkTests, SegmentedTopkLargeSizesStable) TEST(RocprimDeviceSegmentedTopkTests, TopkLargeSizesUnstable) { - // Temporarily disable this test on gfx115x on Windows until we can determine the root cause. - // TODO: remove this after the root cause has been found and fixed properly. -#if defined(_WIN32) - rocprim::detail::target_arch arch = rocprim::detail::target_arch::unknown; - if (rocprim::detail::host_target_arch(hipStreamDefault, arch) != HIP_SUCCESS) - { - std::cerr << "Warning: unable to fetch target architecture for disablement check." << std::endl; - } - - const std::set disabled_arches = { - rocprim::detail::target_arch::gfx1150, - rocprim::detail::target_arch::gfx1151, - rocprim::detail::target_arch::gfx1152, - rocprim::detail::target_arch::gfx1153 - }; - - if (disabled_arches.find(arch) != disabled_arches.end()) - GTEST_SKIP() << "Skipping test on gfx1151 Windows systems due to known issues."; -#endif - constexpr bool descending = false; constexpr bool stable = false; using decomposer_t = rocprim::identity_decomposer; diff --git a/projects/rocprim/test/rocprim/test_device_topk.cpp b/projects/rocprim/test/rocprim/test_device_topk.cpp index 3ec91b809a61..f885ff9f2291 100644 --- a/projects/rocprim/test/rocprim/test_device_topk.cpp +++ b/projects/rocprim/test/rocprim/test_device_topk.cpp @@ -919,26 +919,6 @@ void topk_large_sizes_test(bool debug_synchronous) TEST(RocprimDeviceTopkTests, TopkLargeSizesStable) { - // Temporarily disable this test on gfx115x on Windows until we can determine the root cause. - // TODO: remove this after the root cause has been found and fixed properly. -#if defined(_WIN32) - rocprim::detail::target_arch arch = rocprim::detail::target_arch::unknown; - if (rocprim::detail::host_target_arch(hipStreamDefault, arch) != HIP_SUCCESS) - { - std::cerr << "Warning: unable to fetch target architecture for disablement check." << std::endl; - } - - const std::set disabled_arches = { - rocprim::detail::target_arch::gfx1150, - rocprim::detail::target_arch::gfx1151, - rocprim::detail::target_arch::gfx1152, - rocprim::detail::target_arch::gfx1153 - }; - - if (disabled_arches.find(arch) != disabled_arches.end()) - GTEST_SKIP() << "Skipping test on gfx1151 Windows systems due to known issues."; -#endif - constexpr bool descending = false; constexpr bool stable = true; using decomposer_t = rocprim::identity_decomposer; @@ -963,26 +943,6 @@ TEST(RocprimDeviceTopkTests, TopkLargeSizesStable) TEST(RocprimDeviceTopkTests, TopkLargeSizesUnstable) { - // Temporarily disable this test on gfx115x on Windows until we can determine the root cause. - // TODO: remove this after the root cause has been found and fixed properly. -#if defined(_WIN32) - rocprim::detail::target_arch arch = rocprim::detail::target_arch::unknown; - if (rocprim::detail::host_target_arch(hipStreamDefault, arch) != HIP_SUCCESS) - { - std::cerr << "Warning: unable to fetch target architecture for disablement check." << std::endl; - } - - const std::set disabled_arches = { - rocprim::detail::target_arch::gfx1150, - rocprim::detail::target_arch::gfx1151, - rocprim::detail::target_arch::gfx1152, - rocprim::detail::target_arch::gfx1153 - }; - - if (disabled_arches.find(arch) != disabled_arches.end()) - GTEST_SKIP() << "Skipping test on gfx1151 Windows systems due to known issues."; -#endif - constexpr bool descending = false; constexpr bool stable = false; using decomposer_t = rocprim::identity_decomposer; diff --git a/projects/rocthrust/.clang-format b/projects/rocthrust/.clang-format index 676f9822da5f..ba9b78224095 100644 --- a/projects/rocthrust/.clang-format +++ b/projects/rocthrust/.clang-format @@ -30,7 +30,7 @@ AttributeMacros: [ 'THRUST_HOST', 'THRUST_NODISCARD', 'THRUST_DEPRECATED', - 'THRUST_GLOBAL_CONSTANT', + 'THRUST_INLINE_CONSTANT', # thrust '_CCCL_ALIGNAS_TYPE', '_CCCL_ALIGNAS', diff --git a/projects/rocthrust/CHANGELOG.md b/projects/rocthrust/CHANGELOG.md index 1e3caeb39ac1..f77dc955cca4 100644 --- a/projects/rocthrust/CHANGELOG.md +++ b/projects/rocthrust/CHANGELOG.md @@ -3,15 +3,6 @@ Documentation for rocThrust available at [https://rocm.docs.amd.com/projects/rocThrust/en/latest/](https://rocm.docs.amd.com/projects/rocThrust/en/latest/). -## rocThrust 5.0.0 for ROCm 10.0 - -### Added - -* Largely in feature parity with CCCL/thrust v3.0.3. - - `thrust::tuple`, `thrust::pair` and `thrust::zip_iterator` fall back to rocThrust 4.4.0 implementations when a libhipcxx counterpart corresponding to CCCL/libcudacxx >= v3.0.3 is unavailable, ie - * `thrust::tuple` and `thrust::pair`: some features may differ from CCCL/thrust v3.0.3. - * `thrust::zip_iterator`: some iterator concepts present in CCCL/thrust v3.0.3 are missing. - ## rocThrust 4.5.0 for ROCm 7.14 ### Added diff --git a/projects/rocthrust/CMakeLists.txt b/projects/rocthrust/CMakeLists.txt index 9bdb858bfd21..467bb97f7549 100644 --- a/projects/rocthrust/CMakeLists.txt +++ b/projects/rocthrust/CMakeLists.txt @@ -8,11 +8,11 @@ cmake_minimum_required(VERSION 3.18 FATAL_ERROR) # Update these variables at release time # # Set the library version -set(VERSION_STRING "5.0.0") +set(VERSION_STRING "4.5.0") # Set the minimum required rocPRIM version -set(MIN_ROCPRIM_PACKAGE_VERSION "4.1.0" CACHE STRING "Minimum version of rocPRIM to search for when ROCPRIM_FETCH_METHOD is set to PACKAGE.") +set(MIN_ROCPRIM_PACKAGE_VERSION "4.0.0" CACHE STRING "Minimum version of rocPRIM to search for when ROCPRIM_FETCH_METHOD is set to PACKAGE.") # Set the minimum required rocRAND version -set(MIN_ROCRAND_PACKAGE_VERSION "4.1.0" CACHE STRING "Minimum version of rocRAND to search for when ROCRAND_FETCH_METHOD is set to PACKAGE.") +set(MIN_ROCRAND_PACKAGE_VERSION "4.0.0" CACHE STRING "Minimum version of rocRAND to search for when ROCRAND_FETCH_METHOD is set to PACKAGE.") # Set download branch for dependencies rocPRIM and rocRAND set(ROCM_DEP_RELEASE_BRANCH "release/therock-7.14" CACHE STRING "Download branch for ROCm dependencies") # -------------------------------------- @@ -35,9 +35,10 @@ endif() set(CMAKE_CXX_STANDARD_REQUIRED ON) set(CMAKE_CXX_EXTENSIONS OFF) -set(ALLOWED_CXX_STANDARDS 17 20 23 26) -if(NOT CMAKE_CXX_STANDARD IN_LIST ALLOWED_CXX_STANDARDS) - message(FATAL_ERROR "Only C++17, C++20, C++23 and C++26 are supported") +if (CMAKE_CXX_STANDARD EQUAL 14) + message(WARNING "C++14 will be deprecated in the next major release") +elseif(NOT CMAKE_CXX_STANDARD EQUAL 17) + message(FATAL_ERROR "Only C++14 and C++17 are supported") endif() # Set HIP flags @@ -201,6 +202,10 @@ if(NOT DISABLE_WERROR) list(APPEND COMPILE_OPTIONS -Werror) endif() +if(NOT CMAKE_CXX_STANDARD EQUAL 17) + message(FATAL_ERROR "Only C++17 is supported") +endif() + if (WIN32) add_compile_options(-xhip) add_compile_definitions(THRUST_IGNORE_DEPRECATED_CPP_DIALECT) diff --git a/projects/rocthrust/benchmark/bench/adjacent_difference/basic.cu b/projects/rocthrust/benchmark/bench/adjacent_difference/basic.cu index c0fcecdf658f..33f8f023449b 100644 --- a/projects/rocthrust/benchmark/bench/adjacent_difference/basic.cu +++ b/projects/rocthrust/benchmark/bench/adjacent_difference/basic.cu @@ -1,6 +1,6 @@ /****************************************************************************** * Copyright (c) 2011-2023, NVIDIA CORPORATION. All rights reserved. - * Modifications Copyright (c) 2024-2026, Advanced Micro Devices, Inc. All rights reserved. + * Modifications Copyright (c) 2024-2025, Advanced Micro Devices, Inc. All rights reserved. * * Redistribution and use in source and binary forms, with or without * modification, are permitted provided that the following conditions are met: @@ -110,9 +110,10 @@ void run_benchmark(benchmark::State& state, const std::size_t elements, const st } template -void add_benchmarks(const std::string& name, std::vector& benchmarks, const std::string seed_type) +void add_benchmarks( + const std::string& name, std::vector& benchmarks, const std::string seed_type) { - std::vector bs; + std::vector bs; BENCHMARK_TYPE(int8_t) BENCHMARK_TYPE(int16_t) BENCHMARK_TYPE(int32_t) @@ -139,7 +140,7 @@ int main(int argc, char* argv[]) benchmark::AddCustomContext("seed", seed_type); // Add benchmark - std::vector benchmarks; + std::vector benchmarks; add_benchmarks("basic", benchmarks, seed_type); // Use manual timing diff --git a/projects/rocthrust/benchmark/bench/adjacent_difference/custom.cu b/projects/rocthrust/benchmark/bench/adjacent_difference/custom.cu index 8626c923135e..2be6dc85c80e 100644 --- a/projects/rocthrust/benchmark/bench/adjacent_difference/custom.cu +++ b/projects/rocthrust/benchmark/bench/adjacent_difference/custom.cu @@ -1,6 +1,6 @@ /****************************************************************************** * Copyright (c) 2011-2023, NVIDIA CORPORATION. All rights reserved. - * Modifications Copyright (c) 2024-2026, Advanced Micro Devices, Inc. All rights reserved. + * Modifications Copyright (c) 2024-2025, Advanced Micro Devices, Inc. All rights reserved. * * Redistribution and use in source and binary forms, with or without * modification, are permitted provided that the following conditions are met: @@ -128,9 +128,10 @@ void run_benchmark(benchmark::State& state, const std::size_t elements, const st } template