diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 4102593b0fdd..fb38bff6e868 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -112,6 +112,12 @@ repos: entry: 'rmm::exec_policy\(' language: pygrep types_or: [c, c++, cuda] + - id: use-cuda-iterators + name: use-cuda-iterators + description: 'Enforce that CUDA iterators are used instead of equivalent Thrust iterators' + entry: 'thrust::(?:make_)?(?:constant|counting|discard|permutation|reverse|strided|tabulate_output|transform|transform_input_output|transform_output|zip)_iterator' + language: pygrep + types_or: [c, c++, cuda] - id: exec-policy-nosync-memory-resource name: exec-policy-nosync-memory-resource description: 'Enforce that rmm::exec_policy_nosync is always passed a memory resource as its second argument' diff --git a/cpp/benchmarks/common/generate_input.cu b/cpp/benchmarks/common/generate_input.cu index 2bf1bbd57d28..36cdb711768a 100644 --- a/cpp/benchmarks/common/generate_input.cu +++ b/cpp/benchmarks/common/generate_input.cu @@ -42,7 +42,6 @@ #include #include #include -#include #include #include #include @@ -419,7 +418,7 @@ rmm::device_uvector sample_indices_with_run_length(cudf::size_t thrust::device, run_lens.begin(), run_lens.end(), run_lens.begin(), cuda::std::plus{}); auto const samples_indices = sample_dist(engine, approx_run_len + 1); // This is gather. - auto avg_repeated_sample_indices_iterator = thrust::make_transform_iterator( + auto avg_repeated_sample_indices_iterator = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [rb = run_lens.begin(), @@ -441,14 +440,6 @@ rmm::device_uvector sample_indices_with_run_length(cudf::size_t } } -struct valid_or_zero { - template - __device__ T operator()(cuda::std::tuple len_valid) const - { - return cuda::std::get<1>(len_valid) ? cuda::std::get<0>(len_valid) : T{0}; - } -}; - enum class string_encoding { ASCII, UTF8, @@ -511,13 +502,9 @@ std::unique_ptr create_random_utf8_string_column(data_profile cons lengths.begin(), cuda::proclaim_return_type([] __device__(auto) { return 0; }), cuda::std::logical_not{}); - auto valid_lengths = thrust::make_transform_iterator( - cuda::make_zip_iterator(cuda::std::make_tuple(lengths.begin(), null_mask.begin())), - valid_or_zero{}); - // offsets are created as INT32 or INT64 as appropriate auto [offsets, chars_length] = cudf::strings::detail::make_offsets_child_column( - valid_lengths, valid_lengths + num_rows, stream, mr); + lengths.begin(), lengths.begin() + num_rows, stream, mr); // use the offsetalator to normalize the offset values for use by the string_generator auto offsets_itr = cudf::detail::offsetalator_factory::make_input_iterator(offsets->view()); rmm::device_uvector chars(chars_length, cudf::get_default_stream()); diff --git a/cpp/benchmarks/io/orc/orc_writer_chunks.cpp b/cpp/benchmarks/io/orc/orc_writer_chunks.cpp index 11530238d3da..5be967140498 100644 --- a/cpp/benchmarks/io/orc/orc_writer_chunks.cpp +++ b/cpp/benchmarks/io/orc/orc_writer_chunks.cpp @@ -13,7 +13,7 @@ #include #include -#include +#include #include @@ -92,10 +92,10 @@ void nvbench_orc_chunked_write(nvbench::state& state) auto mem_stats_logger = cudf::memory_stats_logger(); - auto size_iter = thrust::make_transform_iterator( + auto size_iter = cuda::transform_iterator( tables.begin(), [](auto const& i) { return i->num_columns() * i->num_rows(); }); auto row_count_iter = - thrust::make_transform_iterator(tables.begin(), [](auto const& i) { return i->num_rows(); }); + cuda::transform_iterator(tables.begin(), [](auto const& i) { return i->num_rows(); }); auto total_elements = std::accumulate(size_iter, size_iter + num_tables, 0); auto total_rows = std::accumulate(row_count_iter, row_count_iter + num_tables, 0); diff --git a/cpp/benchmarks/iterator/iterator.cu b/cpp/benchmarks/iterator/iterator.cu index 5965776e6469..ead9ca942516 100644 --- a/cpp/benchmarks/iterator/iterator.cu +++ b/cpp/benchmarks/iterator/iterator.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -16,7 +16,6 @@ #include #include #include -#include #include #include diff --git a/cpp/benchmarks/transpose/transpose.cpp b/cpp/benchmarks/transpose/transpose.cpp index beb88f26e12d..8282be0d2d8b 100644 --- a/cpp/benchmarks/transpose/transpose.cpp +++ b/cpp/benchmarks/transpose/transpose.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,7 +12,6 @@ #include #include -#include #include @@ -22,7 +21,7 @@ void bench_transpose(nvbench::state& state) constexpr auto column_type_id = cudf::type_id::INT32; auto int_column_generator = - thrust::make_transform_iterator(cuda::counting_iterator{0}, [count](int i) { + cuda::transform_iterator(cuda::counting_iterator{0}, [count](int i) { return cudf::make_numeric_column( cudf::data_type{column_type_id}, count, cudf::mask_state::ALL_VALID); }); diff --git a/cpp/include/cudf/column/column_device_view.cuh b/cpp/include/cudf/column/column_device_view.cuh index a3b1901244c9..b4238d045b9a 100644 --- a/cpp/include/cudf/column/column_device_view.cuh +++ b/cpp/include/cudf/column/column_device_view.cuh @@ -20,7 +20,6 @@ #include #include #include -#include #include @@ -228,7 +227,7 @@ class alignas(16) column_device_view : public column_device_view_core { * @brief Iterator for navigating this column */ template - using const_iterator = thrust::transform_iterator, count_it>; + using const_iterator = cuda::transform_iterator, count_it>; /** * @brief Return an iterator to the first element of the column. @@ -276,14 +275,14 @@ class alignas(16) column_device_view : public column_device_view_core { */ template using const_optional_iterator = - thrust::transform_iterator, count_it>; + cuda::transform_iterator, count_it>; /** * @brief Pair iterator for navigating this column */ template using const_pair_iterator = - thrust::transform_iterator, count_it>; + cuda::transform_iterator, count_it>; /** * @brief Pair rep iterator for navigating this column @@ -292,7 +291,7 @@ class alignas(16) column_device_view : public column_device_view_core { */ template using const_pair_rep_iterator = - thrust::transform_iterator, count_it>; + cuda::transform_iterator, count_it>; /** * @brief Return an optional iterator to the first element of the column. @@ -701,7 +700,7 @@ class alignas(16) mutable_column_device_view : public mutable_column_device_view * @brief Iterator for navigating this column */ template - using iterator = thrust::transform_iterator, count_it>; + using iterator = cuda::transform_iterator, count_it>; /** * @brief Return first element (accounting for offset) after underlying data diff --git a/cpp/include/cudf/detail/indexalator.cuh b/cpp/include/cudf/detail/indexalator.cuh index ad085f976a03..55b0b483c1fd 100644 --- a/cpp/include/cudf/detail/indexalator.cuh +++ b/cpp/include/cudf/detail/indexalator.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -16,7 +16,6 @@ #include #include #include -#include namespace cudf { namespace detail { @@ -344,8 +343,8 @@ struct indexalator_factory { */ static auto make_input_optional_iterator(scalar const& input, rmm::cuda_stream_view stream) { - return thrust::make_transform_iterator(cuda::make_constant_iterator(0), - scalar_optional_index_accessor{input, stream}); + return cuda::transform_iterator(cuda::make_constant_iterator(0), + scalar_optional_index_accessor{input, stream}); } }; diff --git a/cpp/include/cudf/detail/iterator.cuh b/cpp/include/cudf/detail/iterator.cuh index f9e76dfe0ca9..797bc0acada8 100644 --- a/cpp/include/cudf/detail/iterator.cuh +++ b/cpp/include/cudf/detail/iterator.cuh @@ -30,12 +30,11 @@ #include #include #include -#include namespace cudf { namespace detail { /** - * @brief Convenience wrapper for creating a `thrust::transform_iterator` over a + * @brief Convenience wrapper for creating a `cuda::transform_iterator` over a * `cuda::counting_iterator` within the range [0, INT_MAX]. * * @@ -65,7 +64,7 @@ CUDF_HOST_DEVICE inline auto make_counting_transform_iterator(CountingIterType s cuda::std::numeric_limits::digits, "The `start` for the counting_transform_iterator must be size_type or smaller type"); - return thrust::make_transform_iterator(cuda::counting_iterator{start}, f); + return cuda::transform_iterator(cuda::counting_iterator{start}, f); } /** @@ -381,8 +380,8 @@ auto inline make_scalar_iterator(scalar const& scalar_value) { CUDF_EXPECTS(data_type(type_to_id()) == scalar_value.type(), "the data type mismatch"); CUDF_EXPECTS(scalar_value.is_valid(), "the scalar value must be valid"); - return thrust::make_transform_iterator(cuda::make_constant_iterator(0), - scalar_value_accessor{scalar_value}); + return cuda::transform_iterator(cuda::make_constant_iterator(0), + scalar_value_accessor{scalar_value}); } /** @@ -577,7 +576,7 @@ auto inline make_optional_iterator(scalar const& scalar_value, Nullate has_nulls { CUDF_EXPECTS(type_id_matches_device_storage_type(scalar_value.type().id()), "the data type mismatch"); - return thrust::make_transform_iterator( + return cuda::transform_iterator( cuda::make_constant_iterator(0), scalar_optional_accessor{scalar_value, has_nulls}); } @@ -608,8 +607,8 @@ auto inline make_pair_iterator(scalar const& scalar_value) { CUDF_EXPECTS(type_id_matches_device_storage_type(scalar_value.type().id()), "the data type mismatch"); - return thrust::make_transform_iterator(cuda::make_constant_iterator(0), - scalar_pair_accessor{scalar_value}); + return cuda::transform_iterator(cuda::make_constant_iterator(0), + scalar_pair_accessor{scalar_value}); } /** diff --git a/cpp/include/cudf/detail/join/filtered_join.cuh b/cpp/include/cudf/detail/join/filtered_join.cuh index b6d9949e187b..8bc3c1c19434 100644 --- a/cpp/include/cudf/detail/join/filtered_join.cuh +++ b/cpp/include/cudf/detail/join/filtered_join.cuh @@ -89,7 +89,7 @@ class filtered_join { struct key_pair_fn { CUDF_HOST_DEVICE constexpr key_pair_fn(Hasher const& hasher) : _hasher{hasher} {} - __device__ __forceinline__ auto operator()(size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()(size_type i) const noexcept { return cuco::pair{_hasher(i), T{i}}; } diff --git a/cpp/include/cudf/detail/null_mask.cuh b/cpp/include/cudf/detail/null_mask.cuh index a80a99db8cda..9c04aed6cce7 100644 --- a/cpp/include/cudf/detail/null_mask.cuh +++ b/cpp/include/cudf/detail/null_mask.cuh @@ -26,7 +26,6 @@ #include #include #include -#include #include #include @@ -536,11 +535,11 @@ rmm::device_uvector segmented_count_bits(bitmask_type const* bitmask, static_cast(std::distance(first_bit_indices_begin, first_bit_indices_end)); rmm::device_uvector d_bit_counts(num_ranges, stream); - auto num_set_bits_in_word = thrust::make_transform_iterator(bitmask, popcount{}); + auto num_set_bits_in_word = cuda::transform_iterator(bitmask, popcount{}); auto first_word_indices = - thrust::make_transform_iterator(first_bit_indices_begin, bit_to_word_index{true}); + cuda::transform_iterator(first_bit_indices_begin, bit_to_word_index{true}); auto last_word_indices = - thrust::make_transform_iterator(last_bit_indices_begin, bit_to_word_index{false}); + cuda::transform_iterator(last_bit_indices_begin, bit_to_word_index{false}); // Allocate temporary memory. size_t temp_storage_bytes{0}; @@ -577,7 +576,7 @@ rmm::device_uvector segmented_count_bits(bitmask_type const* bitmask, // Convert from set bits counts to unset bits by subtracting the number of // set bits from the length of the segment. auto segments_begin = cuda::make_zip_iterator(first_bit_indices_begin, last_bit_indices_begin); - auto segment_length_iterator = thrust::transform_iterator( + auto segment_length_iterator = cuda::transform_iterator( segments_begin, cuda::proclaim_return_type([] __device__(auto const& segment) { auto const begin = cuda::std::get<0>(segment); auto const end = cuda::std::get<1>(segment); @@ -683,10 +682,10 @@ std::vector segmented_count_bits(bitmask_type const* bitmask, make_device_uvector_async(h_indices, stream, cudf::get_current_device_resource_ref()); // Compute the bit counts over each segment. - auto first_bit_indices_begin = thrust::make_transform_iterator( + auto first_bit_indices_begin = cuda::transform_iterator( cuda::counting_iterator{0}, index_alternator{false, d_indices.data()}); auto const first_bit_indices_end = first_bit_indices_begin + num_segments; - auto last_bit_indices_begin = thrust::make_transform_iterator( + auto last_bit_indices_begin = cuda::transform_iterator( cuda::counting_iterator{0}, index_alternator{true, d_indices.data()}); rmm::device_uvector d_bit_counts = cudf::detail::segmented_count_bits(bitmask, @@ -794,7 +793,7 @@ std::pair segmented_null_mask_reduction( { auto const segments_begin = cuda::make_zip_iterator(first_bit_indices_begin, last_bit_indices_begin); - auto const segment_length_iterator = thrust::make_transform_iterator( + auto const segment_length_iterator = cuda::transform_iterator( segments_begin, cuda::proclaim_return_type([] __device__(auto const& segment) { auto const begin = cuda::std::get<0>(segment); auto const end = cuda::std::get<1>(segment); diff --git a/cpp/include/cudf/detail/row_operator/hashing.cuh b/cpp/include/cudf/detail/row_operator/hashing.cuh index e71fc5213483..19612737d0e2 100644 --- a/cpp/include/cudf/detail/row_operator/hashing.cuh +++ b/cpp/include/cudf/detail/row_operator/hashing.cuh @@ -22,9 +22,9 @@ #include #include +#include #include #include -#include #include @@ -126,7 +126,7 @@ class device_row_hasher { auto const init = has_columns ? hasher(_table.column(0)) : _seed; auto const start_col = static_cast(has_columns); - auto it = thrust::make_transform_iterator(_table.begin() + start_col, hasher); + auto it = cuda::transform_iterator(_table.begin() + start_col, hasher); return detail::accumulate( it, it + (_table.num_columns() - start_col), init, [](auto hash, auto h) { return cudf::hashing::detail::hash_combine(hash, h); diff --git a/cpp/include/cudf/detail/scatter.cuh b/cpp/include/cudf/detail/scatter.cuh index a967c51963f0..32e33a9b2f5f 100644 --- a/cpp/include/cudf/detail/scatter.cuh +++ b/cpp/include/cudf/detail/scatter.cuh @@ -28,7 +28,6 @@ #include #include #include -#include #include #include #include @@ -380,9 +379,9 @@ std::unique_ptr scatter(table_view const& source, // Transform negative indices to index + target size. auto updated_scatter_map_begin = - thrust::make_transform_iterator(scatter_map_begin, index_converter{target.num_rows()}); + cuda::transform_iterator(scatter_map_begin, index_converter{target.num_rows()}); auto updated_scatter_map_end = - thrust::make_transform_iterator(scatter_map_end, index_converter{target.num_rows()}); + cuda::transform_iterator(scatter_map_end, index_converter{target.num_rows()}); auto result = std::vector>(target.num_columns()); std::transform(source.begin(), diff --git a/cpp/include/cudf/detail/utilities/batched_memset.hpp b/cpp/include/cudf/detail/utilities/batched_memset.hpp index 77f3af1a1835..66b3fa7b1df7 100644 --- a/cpp/include/cudf/detail/utilities/batched_memset.hpp +++ b/cpp/include/cudf/detail/utilities/batched_memset.hpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -16,7 +16,6 @@ #include #include #include -#include #include namespace CUDF_EXPORT cudf { @@ -43,19 +42,18 @@ void batched_memset(cudf::host_span const> host_buffers, host_buffers, stream, cudf::get_current_device_resource_ref()); // Vector of sizes of all buffer spans - auto sizes = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [buffers = buffers.data()] __device__(std::size_t i) { return buffers[i].size(); })); + auto sizes = cuda::transform_iterator( + buffers.begin(), cuda::proclaim_return_type([] __device__(auto const& buffer) { + return buffer.size(); + })); // Constant iterator to the value to memset auto iter_in = cuda::make_constant_iterator(cuda::make_constant_iterator(value)); // Iterator to each device span pointer - auto iter_out = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [buffers = buffers.data()] __device__(std::size_t i) { return buffers[i].data(); })); + auto iter_out = cuda::transform_iterator( + buffers.begin(), + cuda::proclaim_return_type([] __device__(auto const& buffer) { return buffer.data(); })); std::size_t temp_storage_bytes = 0; auto const num_buffers = host_buffers.size(); diff --git a/cpp/include/cudf/detail/utilities/transform_unary_functions.cuh b/cpp/include/cudf/detail/utilities/transform_unary_functions.cuh index baff8c9fdcfd..646a0a6b9208 100644 --- a/cpp/include/cudf/detail/utilities/transform_unary_functions.cuh +++ b/cpp/include/cudf/detail/utilities/transform_unary_functions.cuh @@ -1,22 +1,22 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ /** - * @brief unary functions for thrust::transform_iterator + * @brief unary functions for cuda::transform_iterator * @file transform_unary_functions.cuh * * These are designed for using as AdaptableUnaryFunction - * for thrust::transform_iterator. + * for cuda::transform_iterator. * For the detail of example cases, * @see iterator.cuh iterator_test.cu */ #pragma once +#include #include -#include namespace cudf { /** diff --git a/cpp/include/cudf/lists/detail/scatter.cuh b/cpp/include/cudf/lists/detail/scatter.cuh index 7003d4e7f82e..48d39f054ee0 100644 --- a/cpp/include/cudf/lists/detail/scatter.cuh +++ b/cpp/include/cudf/lists/detail/scatter.cuh @@ -23,7 +23,6 @@ #include #include -#include #include #include #include @@ -104,7 +103,7 @@ std::unique_ptr scatter_impl(rmm::device_uvector cons auto const target_lists_column_view = lists_column_view(target); // Checks that target is a list column. - auto list_size_begin = thrust::make_transform_iterator( + auto list_size_begin = cuda::transform_iterator( target_vector.begin(), cuda::proclaim_return_type([] __device__(unbound_list_view l) { return l.size(); })); auto offsets_column = std::get<0>(cudf::detail::make_offsets_child_column( diff --git a/cpp/include/cudf/lists/list_device_view.cuh b/cpp/include/cudf/lists/list_device_view.cuh index 894fbbf56d74..635fb4081c3b 100644 --- a/cpp/include/cudf/lists/list_device_view.cuh +++ b/cpp/include/cudf/lists/list_device_view.cuh @@ -12,7 +12,6 @@ #include #include #include -#include namespace CUDF_EXPORT cudf { @@ -144,12 +143,12 @@ class list_device_view { /// const pair iterator for the list template using const_pair_iterator = - thrust::transform_iterator, cuda::counting_iterator>; + cuda::transform_iterator, cuda::counting_iterator>; /// const pair iterator type for the list template using const_pair_rep_iterator = - thrust::transform_iterator, cuda::counting_iterator>; + cuda::transform_iterator, cuda::counting_iterator>; /** * @brief Fetcher for a pair iterator to the first element in the list_device_view. diff --git a/cpp/include/cudf/strings/detail/gather.cuh b/cpp/include/cudf/strings/detail/gather.cuh index f683d28cf9be..77df8f218ea4 100644 --- a/cpp/include/cudf/strings/detail/gather.cuh +++ b/cpp/include/cudf/strings/detail/gather.cuh @@ -23,11 +23,11 @@ #include #include +#include #include #include #include #include -#include #include @@ -234,7 +234,7 @@ std::unique_ptr gather(strings_column_view const& strings, strings.is_empty() ? make_empty_column(type_id::INT32)->view() : strings.offsets(), strings.offset()); - auto sizes_itr = thrust::make_transform_iterator( + auto sizes_itr = cuda::transform_iterator( begin, cuda::proclaim_return_type( [d_strings = *d_strings, d_in_offsets] __device__(size_type idx) { @@ -289,7 +289,7 @@ std::unique_ptr gather(strings_column_view const& strings, d_strings->begin(), d_out_chars, offsets_view, begin, output_count); } else { // Iterator over the character column of input strings to gather - auto in_chars_itr = thrust::make_transform_iterator( + auto in_chars_itr = cuda::transform_iterator( begin, cuda::proclaim_return_type([d_strings = *d_strings] __device__(size_type idx) { if (NullifyOutOfBounds && (idx < 0 || idx >= d_strings.size())) { diff --git a/cpp/include/cudf/strings/detail/scatter.cuh b/cpp/include/cudf/strings/detail/scatter.cuh index a3505e5b4213..026da8568cbf 100644 --- a/cpp/include/cudf/strings/detail/scatter.cuh +++ b/cpp/include/cudf/strings/detail/scatter.cuh @@ -15,9 +15,9 @@ #include #include +#include #include #include -#include #include namespace cudf { @@ -64,7 +64,7 @@ std::unique_ptr scatter(SourceIterator begin, // this ensures empty strings are not mapped to nulls in the make_strings_column function auto const size = cuda::std::distance(begin, end); - auto itr = thrust::make_transform_iterator( + auto itr = cuda::transform_iterator( begin, cuda::proclaim_return_type([] __device__(string_view const sv) { return sv.empty() ? string_view{} : sv; })); diff --git a/cpp/include/cudf/strings/detail/strings_children.cuh b/cpp/include/cudf/strings/detail/strings_children.cuh index e06c48c5e272..1f031099c235 100644 --- a/cpp/include/cudf/strings/detail/strings_children.cuh +++ b/cpp/include/cudf/strings/detail/strings_children.cuh @@ -70,22 +70,22 @@ rmm::device_uvector make_chars_buffer(column_view const& offsets, auto chars_data = rmm::device_uvector(chars_size, stream, mr); auto const d_offsets = cudf::detail::offsetalator_factory::make_input_iterator(offsets); - auto const src_ptrs = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type([begin] __device__(uint32_t idx) { + auto const src_ptrs = cuda::transform_iterator( + begin, cuda::proclaim_return_type([] __device__(auto const& index_pair) { // Due to a bug in cub (https://github.com/NVIDIA/cccl/issues/586), - // we have to use `const_cast` to remove `const` qualifier from the source pointer. - // This should be fine as long as we only read but not write anything to the source. - return reinterpret_cast(const_cast(begin[idx].first)); + // we have to use `const_cast` to remove `const` qualifier from the + // source pointer. This should be fine as long as we only read but + // not write anything to the source. + return reinterpret_cast(const_cast(index_pair.first)); })); - auto const src_sizes = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [begin] __device__(uint32_t idx) { return begin[idx].second; })); - auto const dst_ptrs = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type([offsets = d_offsets, output = chars_data.data()] __device__( - uint32_t idx) { return output + offsets[idx]; })); + auto const src_sizes = cuda::transform_iterator( + begin, cuda::proclaim_return_type([] __device__(auto const& index_pair) { + return index_pair.second; + })); + auto const dst_ptrs = cuda::transform_iterator( + d_offsets, + cuda::proclaim_return_type( + [output = chars_data.data()] __device__(auto offset) { return output + offset; })); size_t temp_storage_bytes = 0; CUDF_CUDA_TRY(cub::DeviceMemcpy::Batched( diff --git a/cpp/include/cudf/strings/detail/strings_column_factories.cuh b/cpp/include/cudf/strings/detail/strings_column_factories.cuh index 69b4901e9a72..ad44a46365cf 100644 --- a/cpp/include/cudf/strings/detail/strings_column_factories.cuh +++ b/cpp/include/cudf/strings/detail/strings_column_factories.cuh @@ -15,10 +15,10 @@ #include #include +#include #include #include #include -#include namespace cudf { namespace strings { @@ -56,7 +56,7 @@ std::unique_ptr make_strings_column(IndexPairIterator begin, cuda::proclaim_return_type([] __device__(string_index_pair item) -> size_type { return (item.first != nullptr ? static_cast(item.second) : size_type{0}); }); - auto offsets_transformer_itr = thrust::make_transform_iterator(begin, offsets_transformer); + auto offsets_transformer_itr = cuda::transform_iterator(begin, offsets_transformer); auto [offsets_column, bytes] = cudf::strings::detail::make_offsets_child_column( offsets_transformer_itr, offsets_transformer_itr + strings_count, stream, mr); diff --git a/cpp/include/cudf_test/column_utilities.hpp b/cpp/include/cudf_test/column_utilities.hpp index 1cbee8bdc645..3450d7ad319c 100644 --- a/cpp/include/cudf_test/column_utilities.hpp +++ b/cpp/include/cudf_test/column_utilities.hpp @@ -18,9 +18,9 @@ #include #include +#include #include #include -#include namespace CUDF_EXPORT cudf { namespace test { diff --git a/cpp/include/cudf_test/column_wrapper.hpp b/cpp/include/cudf_test/column_wrapper.hpp index 62579b196294..e749a086cd6b 100644 --- a/cpp/include/cudf_test/column_wrapper.hpp +++ b/cpp/include/cudf_test/column_wrapper.hpp @@ -33,12 +33,12 @@ #include #include #include -#include #include #include #include #include +#include namespace CUDF_EXPORT cudf { namespace test { @@ -162,10 +162,10 @@ rmm::device_buffer make_elements(InputIterator begin, cudf::memory_resources mr) { static_assert(cudf::is_fixed_width(), "Unexpected non-fixed width type."); - auto transformer = fixed_width_type_converter{}; - auto transform_begin = thrust::make_transform_iterator(begin, transformer); - auto const size = cudf::distance(begin, end); - auto const elements = thrust::host_vector(transform_begin, transform_begin + size); + auto const size = cudf::distance(begin, end); + auto elements = thrust::host_vector(size); + std::transform( + begin, end, elements.begin(), fixed_width_type_converter{}); rmm::device_buffer buffer{elements.data(), size * sizeof(ElementTo), stream, mr.get_output_mr()}; stream.synchronize(); // wait for async H2D before host source is destroyed return buffer; @@ -197,11 +197,10 @@ rmm::device_buffer make_elements(InputIterator begin, rmm::cuda_stream_view stream, cudf::memory_resources mr) { - using RepType = typename ElementTo::rep; - auto transformer = fixed_width_type_converter{}; - auto transform_begin = thrust::make_transform_iterator(begin, transformer); - auto const size = cudf::distance(begin, end); - auto const elements = thrust::host_vector(transform_begin, transform_begin + size); + using RepType = typename ElementTo::rep; + auto const size = cudf::distance(begin, end); + auto elements = thrust::host_vector(size); + std::transform(begin, end, elements.begin(), fixed_width_type_converter{}); rmm::device_buffer buffer{elements.data(), size * sizeof(RepType), stream, mr.get_output_mr()}; stream.synchronize(); // wait for async H2D before host source is destroyed return buffer; @@ -235,10 +234,9 @@ rmm::device_buffer make_elements(InputIterator begin, CUDF_EXPECTS(std::all_of(begin, end, [](ElementFrom v) { return v.scale() == 0; }), "Only zero-scale fixed-point values are supported"); - auto to_rep = [](ElementTo fp) { return fp.value(); }; - auto transformer_begin = thrust::make_transform_iterator(begin, to_rep); - auto const size = cudf::distance(begin, end); - auto const elements = thrust::host_vector(transformer_begin, transformer_begin + size); + auto const size = cudf::distance(begin, end); + auto elements = thrust::host_vector(size); + std::transform(begin, end, elements.begin(), [](ElementTo const fp) { return fp.value(); }); rmm::device_buffer buffer{elements.data(), size * sizeof(RepType), stream, mr.get_output_mr()}; stream.synchronize(); // wait for async H2D before host source is destroyed return buffer; @@ -572,13 +570,17 @@ class fixed_width_column_wrapper : public detail::column_wrapper { rmm::cuda_stream_view stream = cudf::test::get_default_stream(), cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { - auto begin = - thrust::make_transform_iterator(elements.begin(), [](auto const& e) { return e.first; }); - auto end = begin + elements.size(); - auto v = - thrust::make_transform_iterator(elements.begin(), [](auto const& e) { return e.second; }); - wrapped = - fixed_width_column_wrapper(begin, end, v, stream, mr).release(); + auto values = std::vector{}; + auto validity = std::vector{}; + values.reserve(elements.size()); + validity.reserve(elements.size()); + for (auto const& [value, valid] : elements) { + values.push_back(value); + validity.push_back(valid); + } + wrapped = fixed_width_column_wrapper( + values.begin(), values.end(), validity.begin(), stream, mr) + .release(); } }; @@ -1009,12 +1011,16 @@ class strings_column_wrapper : public detail::column_wrapper { rmm::cuda_stream_view stream = cudf::test::get_default_stream(), cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { - auto begin = - thrust::make_transform_iterator(strings.begin(), [](auto const& s) { return s.first; }); - auto end = begin + strings.size(); - auto v = - thrust::make_transform_iterator(strings.begin(), [](auto const& s) { return s.second; }); - wrapped = strings_column_wrapper(begin, end, v, stream, mr).release(); + auto values = std::vector{}; + auto validity = std::vector{}; + values.reserve(strings.size()); + validity.reserve(strings.size()); + for (auto const& [value, valid] : strings) { + values.push_back(value); + validity.push_back(valid); + } + wrapped = + strings_column_wrapper(values.begin(), values.end(), validity.begin(), stream, mr).release(); } }; diff --git a/cpp/include/cudf_test/iterator_utilities.hpp b/cpp/include/cudf_test/iterator_utilities.hpp index 22cd5196f6ae..38a3182c3506 100644 --- a/cpp/include/cudf_test/iterator_utilities.hpp +++ b/cpp/include/cudf_test/iterator_utilities.hpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,13 +10,16 @@ #include #include -#include +#include #include +#include +#include namespace CUDF_EXPORT cudf { namespace test { namespace iterators { + /** * @brief Bool iterator for marking (possibly multiple) null elements in a column_wrapper. * @@ -43,9 +46,10 @@ template { using index_type = typename std::iterator_traits::value_type; - return cudf::detail::make_counting_transform_iterator( - 0, [indices = std::vector{index_start, index_end}](auto i) { - return std::find(indices.cbegin(), indices.cend(), i) == indices.cend(); + auto indices = std::make_shared const>(index_start, index_end); + return cuda::transform_iterator( + cuda::counting_iterator{0}, [indices = std::move(indices)](auto i) { + return std::find(indices->cbegin(), indices->cend(), i) == indices->cend(); }); } @@ -165,7 +169,7 @@ template template [[maybe_unused]] static auto nulls_from_nullptrs(std::vector const& ptrs) { - return thrust::make_transform_iterator(ptrs.begin(), [](auto ptr) { return ptr != nullptr; }); + return cuda::transform_iterator(ptrs.begin(), [](auto ptr) { return ptr != nullptr; }); } } // namespace iterators diff --git a/cpp/include/cudf_test/print_utilities.cuh b/cpp/include/cudf_test/print_utilities.cuh index b614d07c3719..484115f308e7 100644 --- a/cpp/include/cudf_test/print_utilities.cuh +++ b/cpp/include/cudf_test/print_utilities.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,8 +11,9 @@ #include -#include +#include +#include #include namespace CUDF_EXPORT cudf { @@ -55,7 +56,7 @@ auto hex(InItT it) { using value_t = typename std::iterator_traits::value_type; using tagged_t = hex_t; - return thrust::make_transform_iterator(it, ToTaggedType{}); + return cuda::transform_iterator(it, ToTaggedType{}); } template && std::is_signed_v)> diff --git a/cpp/src/ast/expression_parser.cpp b/cpp/src/ast/expression_parser.cpp index 8392779e388d..854df2d611f0 100644 --- a/cpp/src/ast/expression_parser.cpp +++ b/cpp/src/ast/expression_parser.cpp @@ -10,7 +10,7 @@ #include #include -#include +#include #include #include @@ -223,7 +223,7 @@ cudf::size_type expression_parser::visit(operation const& expr) auto const operand_data_ref_indices = visit_operands(expr.get_operands()); // Resolve operand types auto data_ref = [this](auto const& index) { return _data_references[index].data_type; }; - auto begin = thrust::make_transform_iterator(operand_data_ref_indices.cbegin(), data_ref); + auto begin = cuda::transform_iterator(operand_data_ref_indices.cbegin(), data_ref); auto end = begin + operand_data_ref_indices.size(); auto const operand_types = std::vector(begin, end); diff --git a/cpp/src/column/column.cu b/cpp/src/column/column.cu index 37c223fba0b8..9cb305fca0cb 100644 --- a/cpp/src/column/column.cu +++ b/cpp/src/column/column.cu @@ -26,8 +26,8 @@ #include +#include #include -#include #include #include @@ -197,7 +197,7 @@ struct create_column_from_view { requires(cudf::is_fixed_width()) { auto op = [&](auto const& child) { return std::make_unique(child, stream, mr); }; - auto begin = thrust::make_transform_iterator(view.child_begin(), op); + auto begin = cuda::transform_iterator(view.child_begin(), op); auto children = std::vector>(begin, begin + view.num_children()); return std::make_unique( diff --git a/cpp/src/column/column_view.cpp b/cpp/src/column/column_view.cpp index b9973b4a2eab..7e11821ffc0a 100644 --- a/cpp/src/column/column_view.cpp +++ b/cpp/src/column/column_view.cpp @@ -14,7 +14,7 @@ #include #include -#include +#include #include #include @@ -214,7 +214,7 @@ void const* mutable_column_view::get_data() const noexcept size_type count_descendants(column_view parent) { auto descendants = [](auto const& child) { return count_descendants(child); }; - auto begin = thrust::make_transform_iterator(parent.child_begin(), descendants); + auto begin = cuda::transform_iterator(parent.child_begin(), descendants); return std::accumulate(begin, begin + parent.num_children(), size_type{parent.num_children()}); } diff --git a/cpp/src/copying/contiguous_split.cu b/cpp/src/copying/contiguous_split.cu index 8fe81ff0a719..27c71b628b7c 100644 --- a/cpp/src/copying/contiguous_split.cu +++ b/cpp/src/copying/contiguous_split.cu @@ -32,7 +32,6 @@ #include #include #include -#include #include #include #include @@ -437,7 +436,7 @@ OutputIter setup_src_buf_data(InputIter begin, InputIter end, OutputIter out_buf template size_type count_src_bufs(InputIter begin, InputIter end) { - auto buf_iter = thrust::make_transform_iterator(begin, [](column_view const& col) { + auto buf_iter = cuda::transform_iterator(begin, [](column_view const& col) { auto const children_counts = count_src_bufs(col.child_begin(), col.child_end()); return 1 + (col.nullable() ? 1 : 0) + children_counts; }); @@ -1876,7 +1875,7 @@ struct contiguous_split_state { auto const keys = cudf::detail::make_counting_transform_iterator( 0, out_to_in_index_function{chunk_iter_state->d_batch_offsets.begin(), (int)num_bufs}); - auto values = thrust::make_transform_iterator( + auto values = cuda::transform_iterator( chunk_iter_state->d_batched_dst_buf_info.begin(), cuda::proclaim_return_type( [] __device__(dst_buf_info const& info) { return info.valid_count; })); diff --git a/cpp/src/copying/gather.cu b/cpp/src/copying/gather.cu index 0e84a51261d8..055003e01808 100644 --- a/cpp/src/copying/gather.cu +++ b/cpp/src/copying/gather.cu @@ -15,8 +15,8 @@ #include #include +#include #include -#include #include @@ -41,8 +41,8 @@ std::unique_ptr
gather(table_view const& source_table, auto idx_converter = cuda::proclaim_return_type( [n_rows] __device__(size_type in) { return in < 0 ? in + n_rows : in; }); return gather(source_table, - thrust::make_transform_iterator(map_begin, idx_converter), - thrust::make_transform_iterator(map_end, idx_converter), + cuda::transform_iterator(map_begin, idx_converter), + cuda::transform_iterator(map_end, idx_converter), bounds_policy, stream, mr); diff --git a/cpp/src/copying/scatter.cu b/cpp/src/copying/scatter.cu index bb2df3867e96..9fbeabd96135 100644 --- a/cpp/src/copying/scatter.cu +++ b/cpp/src/copying/scatter.cu @@ -30,7 +30,6 @@ #include #include #include -#include #include #include @@ -348,7 +347,7 @@ std::unique_ptr
scatter(std::vector> // note: the intermediate ((in % n_rows) + n_rows) will overflow a size_type for any value of `in` // > (2^31)/2, but the end result after the final (% n_rows) will fit. so we'll do the computation // using a signed 64 bit value. - auto scatter_iter = thrust::make_transform_iterator( + auto scatter_iter = cuda::transform_iterator( map_begin, cuda::proclaim_return_type( [n_rows = static_cast(n_rows)] __device__(size_type in) -> size_type { diff --git a/cpp/src/copying/segmented_shift.cu b/cpp/src/copying/segmented_shift.cu index e3453395d688..99df9ac15dd1 100644 --- a/cpp/src/copying/segmented_shift.cu +++ b/cpp/src/copying/segmented_shift.cu @@ -13,10 +13,10 @@ #include #include +#include #include #include #include -#include namespace cudf { namespace detail { diff --git a/cpp/src/copying/slice.cu b/cpp/src/copying/slice.cu index ffdfa31468a9..c76f3e658ecf 100644 --- a/cpp/src/copying/slice.cu +++ b/cpp/src/copying/slice.cu @@ -13,8 +13,8 @@ #include #include +#include #include -#include #include #include @@ -97,7 +97,7 @@ std::vector slice(table_view const& input, // 2d arrangement of column_views that represent the outgoing table_views sliced_table[i][j] // where i is the i'th column of the j'th table_view auto op = [&indices, &stream](auto const& c) { return cudf::detail::slice(c, indices, stream); }; - auto f = thrust::make_transform_iterator(input.begin(), op); + auto f = cuda::transform_iterator(input.begin(), op); auto sliced_table = std::vector>(f, f + input.num_columns()); sliced_table.reserve(indices.size() + 1); diff --git a/cpp/src/groupby/sort/group_count.cu b/cpp/src/groupby/sort/group_count.cu index c697c76d774b..a527190c10d0 100644 --- a/cpp/src/groupby/sort/group_count.cu +++ b/cpp/src/groupby/sort/group_count.cu @@ -17,7 +17,6 @@ #include #include #include -#include namespace cudf { namespace groupby { @@ -43,10 +42,9 @@ std::unique_ptr group_count_valid(column_view const& values, // make_validity_iterator returns a boolean iterator that sums to 1 (1+1=1) // so we need to transform it to cast it to an integer type auto bitmask_iterator = - thrust::make_transform_iterator(cudf::detail::make_validity_iterator(*values_view), - cuda::proclaim_return_type([] __device__(auto b) { - return static_cast(b); - })); + cuda::transform_iterator(cudf::detail::make_validity_iterator(*values_view), + cuda::proclaim_return_type( + [] __device__(auto b) { return static_cast(b); })); cudf::detail::reduce_by_key_async(group_labels.begin(), group_labels.end(), diff --git a/cpp/src/groupby/sort/group_nth_element.cu b/cpp/src/groupby/sort/group_nth_element.cu index ee992e9f29b8..40f438c042d6 100644 --- a/cpp/src/groupby/sort/group_nth_element.cu +++ b/cpp/src/groupby/sort/group_nth_element.cu @@ -20,7 +20,6 @@ #include #include #include -#include #include #include #include @@ -72,10 +71,9 @@ std::unique_ptr group_nth_element(column_view const& values, // Returns index of nth value. auto values_view = column_device_view::create(values, stream); auto bitmask_iterator = - thrust::make_transform_iterator(cudf::detail::make_validity_iterator(*values_view), - cuda::proclaim_return_type([] __device__(auto b) { - return static_cast(b); - })); + cuda::transform_iterator(cudf::detail::make_validity_iterator(*values_view), + cuda::proclaim_return_type( + [] __device__(auto b) { return static_cast(b); })); rmm::device_uvector intra_group_index(values.size(), stream); // intra group index for valids only. thrust::exclusive_scan_by_key( diff --git a/cpp/src/groupby/sort/group_nunique.cu b/cpp/src/groupby/sort/group_nunique.cu index e3a54734c6d9..f99ba0a88739 100644 --- a/cpp/src/groupby/sort/group_nunique.cu +++ b/cpp/src/groupby/sort/group_nunique.cu @@ -15,7 +15,6 @@ #include #include -#include namespace cudf { namespace groupby { diff --git a/cpp/src/groupby/sort/group_scan_util.cuh b/cpp/src/groupby/sort/group_scan_util.cuh index 961e124ae9f4..32c1d9475542 100644 --- a/cpp/src/groupby/sort/group_scan_util.cuh +++ b/cpp/src/groupby/sort/group_scan_util.cuh @@ -28,7 +28,6 @@ #include #include #include -#include #include namespace cudf { @@ -118,14 +117,14 @@ struct group_scan_functor() }; if (values.has_nulls()) { - auto input = thrust::make_transform_iterator( + auto input = cuda::transform_iterator( make_null_replacement_iterator(*values_view, OpType::template identity()), cudf::detail::cast_fn{}); do_scan(input, result_view->begin(), OpType{}); result->set_null_mask(cudf::detail::copy_bitmask(values, stream, mr), values.null_count()); } else { - auto input = thrust::make_transform_iterator(values_view->begin(), - cudf::detail::cast_fn{}); + auto input = cuda::transform_iterator(values_view->begin(), + cudf::detail::cast_fn{}); do_scan(input, result_view->begin(), OpType{}); } return result; diff --git a/cpp/src/groupby/sort/group_std.cu b/cpp/src/groupby/sort/group_std.cu index e75069b97481..35341373df29 100644 --- a/cpp/src/groupby/sort/group_std.cu +++ b/cpp/src/groupby/sort/group_std.cu @@ -23,7 +23,6 @@ #include #include #include -#include #include namespace cudf { diff --git a/cpp/src/io/csv/durations.cu b/cpp/src/io/csv/durations.cu index da150552e5a1..f4ae0a7aa50f 100644 --- a/cpp/src/io/csv/durations.cu +++ b/cpp/src/io/csv/durations.cu @@ -17,7 +17,6 @@ #include #include #include -#include namespace cudf { namespace io { diff --git a/cpp/src/io/csv/reader_impl.cu b/cpp/src/io/csv/reader_impl.cu index 1078a69e4bdb..4d14c071d185 100644 --- a/cpp/src/io/csv/reader_impl.cu +++ b/cpp/src/io/csv/reader_impl.cu @@ -1025,17 +1025,14 @@ table_with_metadata read_csv(cudf::io::datasource* source, auto const replaced_all_iter = cudf::detail::make_optional_iterator( *replaced_all_view, cudf::nullate::DYNAMIC{replaced_all_col->nullable()}); - auto const* original_pairs = buffer->_strings->data(); - auto const original_iter = thrust::make_transform_iterator( - cuda::counting_iterator{0}, + auto const original_iter = cuda::transform_iterator( + buffer->_strings->data(), cuda::proclaim_return_type>( - [original_pairs] __device__( - size_type idx) -> cuda::std::optional { - auto const& p = original_pairs[idx]; - return p.first != nullptr - ? cuda::std::optional{cudf::string_view{p.first, - p.second}} - : cuda::std::nullopt; + [] __device__(auto const& pair) -> cuda::std::optional { + return pair.first != nullptr + ? cuda::std::optional{cudf::string_view{pair.first, + pair.second}} + : cuda::std::nullopt; })); out_columns[col_idx] = cudf::strings::detail::copy_if_else( diff --git a/cpp/src/io/fst/logical_stack.cuh b/cpp/src/io/fst/logical_stack.cuh index 054f2845d33b..c4b2ededdd8e 100644 --- a/cpp/src/io/fst/logical_stack.cuh +++ b/cpp/src/io/fst/logical_stack.cuh @@ -17,6 +17,7 @@ #include #include +#include #include #include #include @@ -323,7 +324,7 @@ void sparse_stack_op_to_top_of_stack(StackSymbolItT d_symbols, using StackSymbolToStackOpT = detail::StackSymbolToStackOp; // transform_iterator converting stack symbols to stack operations - using TransformInputItT = thrust::transform_iterator; + using TransformInputItT = cuda::transform_iterator; constexpr bool supports_reset_op = SupportResetOperation == stack_op_support::WITH_RESET_SUPPORT; @@ -354,7 +355,7 @@ void sparse_stack_op_to_top_of_stack(StackSymbolItT d_symbols, // with the empty_stack_symbol StackOpT const empty_stack{0, empty_stack_symbol}; - thrust::transform_iterator, StackOpT*> kv_ops_scan_in( + cuda::transform_iterator, StackOpT*> kv_ops_scan_in( nullptr, detail::RemapEmptyStack{empty_stack}); StackOpT* kv_ops_scan_out = nullptr; @@ -367,7 +368,7 @@ void sparse_stack_op_to_top_of_stack(StackSymbolItT d_symbols, // operation if constexpr (supports_reset_op) { // Iterator that returns `1` for every symbol that corresponds to a `reset` operation - auto reset_segments_it = thrust::make_transform_iterator( + auto reset_segments_it = cuda::transform_iterator( d_symbols, detail::NewlineToResetStackSegmentOp{symbol_to_stack_op}); @@ -468,7 +469,7 @@ void sparse_stack_op_to_top_of_stack(StackSymbolItT d_symbols, // Compute prefix sum of the stack level after each operation if constexpr (supports_reset_op) { // Iterator that returns `1` for every symbol that corresponds to a `reset` operation - auto reset_segments_it = thrust::make_transform_iterator( + auto reset_segments_it = cuda::transform_iterator( d_symbols, detail::NewlineToResetStackSegmentOp{symbol_to_stack_op}); @@ -543,7 +544,7 @@ void sparse_stack_op_to_top_of_stack(StackSymbolItT d_symbols, read_symbol); // transform_iterator the stack operations to the stack symbol they represent - thrust::transform_iterator kv_op_to_stack_sym_it( + cuda::transform_iterator kv_op_to_stack_sym_it( kv_ops_scan_out, detail::StackOpToStackSymbol{}); // Scatter the stack symbols to the output tape (spots that are not scattered to have been diff --git a/cpp/src/io/json/host_tree_algorithms.cu b/cpp/src/io/json/host_tree_algorithms.cu index 6070e01ae490..3d750e5d4338 100644 --- a/cpp/src/io/json/host_tree_algorithms.cu +++ b/cpp/src/io/json/host_tree_algorithms.cu @@ -1068,7 +1068,7 @@ void scatter_offsets(tree_meta_t const& tree, // col[parent_col_id].child_offsets[row_offset[parent_node_id]] auto& parent_col_ids = sorted_col_ids; // reuse sorted_col_ids - auto parent_col_id = thrust::make_transform_iterator( + auto parent_col_id = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [col_ids = col_ids.begin(), diff --git a/cpp/src/io/json/json_normalization.cu b/cpp/src/io/json/json_normalization.cu index 81f6cc140cd9..59fcfb47b3f4 100644 --- a/cpp/src/io/json/json_normalization.cu +++ b/cpp/src/io/json/json_normalization.cu @@ -356,16 +356,14 @@ std:: inbuf_offsets.begin(), 0); - auto input_it = thrust::make_transform_iterator( - cuda::counting_iterator{0}, + auto input_it = cuda::transform_iterator( + col_offsets.begin(), cuda::proclaim_return_type( - [d_input = d_input.begin(), col_offsets = col_offsets.begin()] __device__( - std::size_t i) -> char const* { return &d_input[col_offsets[i]]; })); - auto output_it = thrust::make_transform_iterator( - cuda::counting_iterator{0}, + [d_input = d_input.begin()] __device__(auto offset) { return &d_input[offset]; })); + auto output_it = cuda::transform_iterator( + inbuf_offsets.cbegin(), cuda::proclaim_return_type( - [inbuf = inbuf.begin(), inbuf_offsets = inbuf_offsets.cbegin()] __device__( - std::size_t i) -> char* { return &inbuf[inbuf_offsets[i]]; })); + [inbuf = inbuf.begin()] __device__(auto offset) { return &inbuf[offset]; })); { // cub device batched copy diff --git a/cpp/src/io/json/json_tree.cu b/cpp/src/io/json/json_tree.cu index 273d59651c7e..4c9ff693c573 100644 --- a/cpp/src/io/json/json_tree.cu +++ b/cpp/src/io/json/json_tree.cu @@ -36,8 +36,6 @@ #include #include #include -#include -#include #include #include #include @@ -295,7 +293,7 @@ tree_meta_t get_tree_representation(device_span tokens, rmm::device_uvector node_levels(num_nodes, stream, mr); { rmm::device_uvector token_levels(num_tokens, stream); - auto const push_pop_it = thrust::make_transform_iterator( + auto const push_pop_it = cuda::transform_iterator( tokens.begin(), cuda::proclaim_return_type( [does_push, does_pop] __device__(PdaTokenT const token) -> size_type { @@ -438,7 +436,7 @@ tree_meta_t get_tree_representation(device_span tokens, rmm::device_uvector token_levels(num_nested, stream); rmm::device_uvector token_id(num_nested, stream); rmm::device_uvector parent_node_ids(num_nested, stream); - auto const push_pop_it = thrust::make_transform_iterator( + auto const push_pop_it = cuda::transform_iterator( tokens.begin(), cuda::proclaim_return_type( [] __device__(PdaTokenT const token) -> size_type { @@ -501,20 +499,19 @@ tree_meta_t get_tree_representation(device_span tokens, stream); // scatter to node_range_end for only nested end tokens. - auto token_indices_it = - thrust::make_permutation_iterator(token_indices.begin(), token_id.begin()); - auto nested_node_range_end_it = - thrust::make_transform_output_iterator(node_range_end.begin(), [] __device__(auto i) { + auto token_indices_it = cuda::transform_iterator( + cuda::make_permutation_iterator(token_indices.begin(), token_id.begin()), + [] __device__(auto i) -> SymbolOffsetT { // add +1 to include end symbol. return i + 1; }); - auto stencil = thrust::make_transform_iterator(token_id.begin(), is_nested_end{tokens.data()}); + auto stencil = cuda::transform_iterator(token_id.begin(), is_nested_end{tokens.data()}); thrust::scatter_if(rmm::exec_policy_nosync(stream, cudf::get_current_device_resource_ref()), token_indices_it, token_indices_it + num_nested, parent_node_ids.begin(), stencil, - nested_node_range_end_it); + node_range_end.begin()); } return {std::move(node_categories), diff --git a/cpp/src/io/json/nested_json_gpu.cu b/cpp/src/io/json/nested_json_gpu.cu index 73d800a42881..d2dc75d10c02 100644 --- a/cpp/src/io/json/nested_json_gpu.cu +++ b/cpp/src/io/json/nested_json_gpu.cu @@ -29,7 +29,6 @@ #include #include -#include #include #include diff --git a/cpp/src/io/orc/reader_impl_chunking.cu b/cpp/src/io/orc/reader_impl_chunking.cu index 3c38cc6bb13d..87d93b8de79a 100644 --- a/cpp/src/io/orc/reader_impl_chunking.cu +++ b/cpp/src/io/orc/reader_impl_chunking.cu @@ -19,9 +19,9 @@ #include #include +#include #include #include -#include #include #include @@ -144,10 +144,10 @@ std::vector find_splits(host_span cumulative_sizes, [[maybe_unused]] std::size_t cur_cumulative_rows{0}; - auto const start = thrust::make_transform_iterator( - cumulative_sizes.begin(), - [&](auto const& size) { return size.size_bytes - cur_cumulative_size; }); - auto const end = start + cumulative_sizes.size(); + auto const start = cuda::transform_iterator(cumulative_sizes.begin(), [&](auto const& size) { + return size.size_bytes - cur_cumulative_size; + }); + auto const end = start + cumulative_sizes.size(); while (cur_count < total_count) { int64_t split_pos = static_cast(cuda::std::distance( diff --git a/cpp/src/io/orc/writer_impl.cu b/cpp/src/io/orc/writer_impl.cu index 951c53f780df..b58ee2c28f4f 100644 --- a/cpp/src/io/orc/writer_impl.cu +++ b/cpp/src/io/orc/writer_impl.cu @@ -46,7 +46,6 @@ #include #include #include -#include #include #include #include @@ -1473,8 +1472,8 @@ encoded_footer_statistics finish_statistic_blobs(Footer const& footer, persisted_statistics& per_chunk_stats, cuda::stream_ref stream) { - auto stripe_size_iter = thrust::make_transform_iterator(per_chunk_stats.stripe_stat_merge.begin(), - [](auto const& s) { return s.size(); }); + auto stripe_size_iter = cuda::transform_iterator(per_chunk_stats.stripe_stat_merge.begin(), + [](auto const& s) { return s.size(); }); auto const num_columns = footer.types.size() - 1; auto const num_stripes = footer.stripes.size(); diff --git a/cpp/src/io/parquet/experimental/hybrid_scan_chunking.cu b/cpp/src/io/parquet/experimental/hybrid_scan_chunking.cu index b64b23ba67f7..99728cc00ae1 100644 --- a/cpp/src/io/parquet/experimental/hybrid_scan_chunking.cu +++ b/cpp/src/io/parquet/experimental/hybrid_scan_chunking.cu @@ -148,8 +148,8 @@ void hybrid_scan_reader_impl::setup_next_pass( // store off how much memory we've used so far. This includes the compressed page data and the // decompressed dictionary data. we will subtract this from the available total memory for the // subpasses - auto chunk_iter = thrust::make_transform_iterator(pass.chunks.d_begin(), - parquet::detail::get_chunk_compressed_size{}); + auto chunk_iter = + cuda::transform_iterator(pass.chunks.d_begin(), parquet::detail::get_chunk_compressed_size{}); pass.base_mem_size = decomp_dict_data_size + cudf::detail::reduce( diff --git a/cpp/src/io/parquet/experimental/page_index_filter.cu b/cpp/src/io/parquet/experimental/page_index_filter.cu index 6f63b711f41e..61312327d8ac 100644 --- a/cpp/src/io/parquet/experimental/page_index_filter.cu +++ b/cpp/src/io/parquet/experimental/page_index_filter.cu @@ -250,30 +250,22 @@ struct page_stats_caster : public stats_caster_base { auto row_str_chars = rmm::device_buffer(total_bytes, stream, mr); // Iterator for input (page-level) string chars - auto src_iter = thrust::make_transform_iterator( - cuda::counting_iterator{0}, + auto const page_offsets = + cuda::make_permutation_iterator(page_str_offsets.begin(), page_indices.begin()); + auto src_iter = cuda::transform_iterator( + page_offsets, cuda::proclaim_return_type( - [chars = page_str_chars.begin(), - offsets = page_str_offsets.begin(), - page_indices = page_indices.begin()] __device__(std::size_t index) { - auto const page_index = page_indices[index]; - return chars + offsets[page_index]; - })); + [chars = page_str_chars.begin()] __device__(auto offset) { return chars + offset; })); // Iterator for output (row-level) string chars - auto dst_iter = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [chars = reinterpret_cast(row_str_chars.data()), - offsets = row_str_offsets.begin()] __device__(std::size_t index) { - return chars + offsets[index]; - })); + auto dst_iter = + cuda::transform_iterator(row_str_offsets.begin(), + cuda::proclaim_return_type( + [chars = reinterpret_cast(row_str_chars.data())] __device__( + auto offset) { return chars + offset; })); // Iterator for string sizes - auto size_iter = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [sizes = row_str_sizes.begin()] __device__(std::size_t index) { return sizes[index]; })); + auto size_iter = row_str_sizes.begin(); // Gather page-level string chars to row-level string chars cudf::detail::batched_memcpy_async(src_iter, dst_iter, size_iter, total_rows, stream); diff --git a/cpp/src/io/parquet/page_data.cu b/cpp/src/io/parquet/page_data.cu index ad5d48fbca83..54f43b9585d3 100644 --- a/cpp/src/io/parquet/page_data.cu +++ b/cpp/src/io/parquet/page_data.cu @@ -14,7 +14,6 @@ #include #include -#include namespace cudf::io::parquet::detail { @@ -576,7 +575,7 @@ void write_final_offsets(host_span offsets, auto d_src_data = cudf::detail::make_device_uvector_async( offsets, stream, cudf::get_current_device_resource_ref()); // Iterator for the source (scalar) data - auto src_iter = thrust::make_transform_iterator( + auto src_iter = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [src = d_src_data.begin()] __device__(std::size_t i) { return src + i; })); diff --git a/cpp/src/io/parquet/page_enc.cu b/cpp/src/io/parquet/page_enc.cu index 129e71dea9b2..b7fb422ab356 100644 --- a/cpp/src/io/parquet/page_enc.cu +++ b/cpp/src/io/parquet/page_enc.cu @@ -32,7 +32,6 @@ #include #include #include -#include #include #include #include diff --git a/cpp/src/io/parquet/parquet_gpu.hpp b/cpp/src/io/parquet/parquet_gpu.hpp index 2b1b49a4d376..0f487155b4f1 100644 --- a/cpp/src/io/parquet/parquet_gpu.hpp +++ b/cpp/src/io/parquet/parquet_gpu.hpp @@ -21,6 +21,7 @@ #include #include +#include #include #include #include @@ -408,7 +409,7 @@ struct get_page_key { */ inline auto make_page_key_iterator(device_span pages) { - return thrust::make_transform_iterator(pages.begin(), get_page_key{}); + return cuda::transform_iterator(pages.begin(), get_page_key{}); } /** diff --git a/cpp/src/io/parquet/reader_impl_chunking.cu b/cpp/src/io/parquet/reader_impl_chunking.cu index aef477a31c92..1cb95bd35a97 100644 --- a/cpp/src/io/parquet/reader_impl_chunking.cu +++ b/cpp/src/io/parquet/reader_impl_chunking.cu @@ -167,8 +167,7 @@ void reader_impl::setup_next_pass(read_mode mode) // store off how much memory we've used so far. This includes the compressed page data and the // decompressed dictionary data. we will subtract this from the available total memory for the // subpasses - auto chunk_iter = - thrust::make_transform_iterator(pass.chunks.d_begin(), get_chunk_compressed_size{}); + auto chunk_iter = cuda::transform_iterator(pass.chunks.d_begin(), get_chunk_compressed_size{}); pass.base_mem_size = decomp_dict_data_size + cudf::detail::reduce( @@ -254,7 +253,7 @@ void reader_impl::setup_next_subpass(read_mode mode) // indices rmm::device_uvector c_info(pass.pages.size(), _stream); auto page_keys = make_page_key_iterator(pass.pages); - auto page_size = thrust::make_transform_iterator(pass.pages.d_begin(), get_page_input_size{}); + auto page_size = cuda::transform_iterator(pass.pages.d_begin(), get_page_input_size{}); thrust::inclusive_scan_by_key( rmm::exec_policy_nosync(_stream, cudf::get_current_device_resource_ref()), page_keys, @@ -576,9 +575,8 @@ void reader_impl::compute_output_chunks_for_subpass() // generate row_indices and cumulative output sizes for all pages rmm::device_uvector c_info(subpass.pages.size(), _stream); - auto page_input = - thrust::make_transform_iterator(subpass.pages.device_begin(), get_page_output_size{}); - auto page_keys = make_page_key_iterator(subpass.pages); + auto page_input = cuda::transform_iterator(subpass.pages.device_begin(), get_page_output_size{}); + auto page_keys = make_page_key_iterator(subpass.pages); thrust::inclusive_scan_by_key( rmm::exec_policy_nosync(_stream, cudf::get_current_device_resource_ref()), page_keys, diff --git a/cpp/src/io/parquet/reader_impl_preprocess.cu b/cpp/src/io/parquet/reader_impl_preprocess.cu index 30ed747142eb..fc35c234140f 100644 --- a/cpp/src/io/parquet/reader_impl_preprocess.cu +++ b/cpp/src/io/parquet/reader_impl_preprocess.cu @@ -28,7 +28,6 @@ #include #include #include -#include #include #include @@ -688,8 +687,8 @@ void reader_impl::generate_list_column_row_counts(is_estimate_row_counts is_esti pass.pages.d_begin(), pass.pages.d_end(), set_list_row_count_estimate{pass.chunks}); - auto key_input = thrust::make_transform_iterator(pass.pages.d_begin(), get_page_chunk_idx{}); - auto page_input = thrust::make_transform_iterator(pass.pages.d_begin(), get_page_num_rows{}); + auto key_input = cuda::transform_iterator(pass.pages.d_begin(), get_page_chunk_idx{}); + auto page_input = cuda::transform_iterator(pass.pages.d_begin(), get_page_num_rows{}); thrust::exclusive_scan_by_key( rmm::exec_policy_nosync(_stream, cudf::get_current_device_resource_ref()), key_input, @@ -802,8 +801,8 @@ void reader_impl::preprocess_subpass_pages(read_mode mode, size_t chunk_read_lim // field in ColumnChunkDesc is the absolute row index for the whole file. chunk_row in PageInfo is // relative to the beginning of the chunk. so in the kernels, chunk.start_row + page.chunk_row // gives us the absolute row index - auto key_input = thrust::make_transform_iterator(pass.pages.d_begin(), get_page_chunk_idx{}); - auto page_input = thrust::make_transform_iterator(pass.pages.d_begin(), get_page_num_rows{}); + auto key_input = cuda::transform_iterator(pass.pages.d_begin(), get_page_chunk_idx{}); + auto page_input = cuda::transform_iterator(pass.pages.d_begin(), get_page_num_rows{}); thrust::exclusive_scan_by_key( rmm::exec_policy_nosync(_stream, cudf::get_current_device_resource_ref()), key_input, @@ -1036,7 +1035,7 @@ void reader_impl::allocate_columns(read_mode mode, size_t skip_rows, size_t num_ d_cols_info.data(), max_depth, subpass.pages.size(), subpass.pages.device_begin()}); // Manually create a size_t `key_start` compatible counting_transform_iterator. - auto const reduction_keys = thrust::make_transform_iterator( + auto const reduction_keys = cuda::transform_iterator( cuda::counting_iterator{key_start}, get_reduction_key{subpass.pages.size()}); // Find the size of each column @@ -1139,8 +1138,8 @@ cudf::detail::host_vector reader_impl::calculate_page_string_offsets() rmm::device_uvector d_col_sizes(_input_columns.size(), _stream); // use page_index to fetch page string sizes in the proper order - auto val_iter = thrust::make_transform_iterator(subpass.pages.device_begin(), - page_to_string_size{pass.chunks.d_begin()}); + auto val_iter = cuda::transform_iterator(subpass.pages.device_begin(), + page_to_string_size{pass.chunks.d_begin()}); // do scan by key to calculate string offsets for each page thrust::exclusive_scan_by_key( diff --git a/cpp/src/io/parquet/reader_impl_preprocess_utils.cu b/cpp/src/io/parquet/reader_impl_preprocess_utils.cu index 4cc1750ca61f..5bc9329f33f8 100644 --- a/cpp/src/io/parquet/reader_impl_preprocess_utils.cu +++ b/cpp/src/io/parquet/reader_impl_preprocess_utils.cu @@ -21,7 +21,6 @@ #include #include #include -#include #include #include #include diff --git a/cpp/src/io/utilities/data_casting.cu b/cpp/src/io/utilities/data_casting.cu index 9503e945e32a..11be633608f8 100644 --- a/cpp/src/io/utilities/data_casting.cu +++ b/cpp/src/io/utilities/data_casting.cu @@ -935,7 +935,7 @@ std::unique_ptr parse_data( } // Prepare iterator that returns (string_ptr, string_length)-pairs needed by type conversion - auto str_tuples = thrust::make_transform_iterator(offset_length_begin, to_string_view_pair{data}); + auto str_tuples = cuda::transform_iterator(offset_length_begin, to_string_view_pair{data}); if (col_type == cudf::data_type{cudf::type_id::STRING}) { return parse_string( diff --git a/cpp/src/jit/helpers.cpp b/cpp/src/jit/helpers.cpp index a016ca34dce5..061ec9a4a8d5 100644 --- a/cpp/src/jit/helpers.cpp +++ b/cpp/src/jit/helpers.cpp @@ -9,6 +9,8 @@ #include #include +#include + #include #include #include @@ -54,8 +56,8 @@ size_type get_projection_size(std::span build_ptx_params(std::span output_typenames, diff --git a/cpp/src/jit/helpers.hpp b/cpp/src/jit/helpers.hpp index 9def0042525e..2abc300adeee 100644 --- a/cpp/src/jit/helpers.hpp +++ b/cpp/src/jit/helpers.hpp @@ -10,7 +10,7 @@ #include #include -#include +#include #include #include diff --git a/cpp/src/join/distinct_hash_join.cu b/cpp/src/join/distinct_hash_join.cu index b7d735cfb78d..94043b12413f 100644 --- a/cpp/src/join/distinct_hash_join.cu +++ b/cpp/src/join/distinct_hash_join.cu @@ -52,7 +52,7 @@ class primitive_keys_fn { public: CUDF_HOST_DEVICE constexpr primitive_keys_fn(hasher const& hash) : _hash{hash} {} - __device__ __forceinline__ auto operator()(size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()(size_type i) const noexcept { return cuco::pair{_hash(i), T{i}}; } @@ -72,7 +72,7 @@ class build_keys_fn { public: CUDF_HOST_DEVICE constexpr build_keys_fn(hasher const& hash) : _hash{hash} {} - __device__ __forceinline__ auto operator()(size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()(size_type i) const noexcept { return cuco::pair{_hash(i), T{i}}; } diff --git a/cpp/src/join/key_remapping.cu b/cpp/src/join/key_remapping.cu index 8191ea164e08..8d0787a67831 100644 --- a/cpp/src/join/key_remapping.cu +++ b/cpp/src/join/key_remapping.cu @@ -94,7 +94,8 @@ class make_key_pair { public: CUDF_HOST_DEVICE constexpr make_key_pair(Hasher const& hash) : _hash{hash} {} - __device__ __forceinline__ auto operator()(cudf::size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()( + cudf::size_type i) const noexcept { return cuco::pair{_hash(i), T{i}}; } diff --git a/cpp/src/join/mark_join.cuh b/cpp/src/join/mark_join.cuh index 9c83e7bfa855..3cb8922b7c41 100644 --- a/cpp/src/join/mark_join.cuh +++ b/cpp/src/join/mark_join.cuh @@ -78,7 +78,7 @@ template struct masked_key_fn { CUDF_HOST_DEVICE constexpr masked_key_fn(Hasher const& hasher) : _hasher{hasher} {} - __device__ __forceinline__ auto operator()(size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()(size_type i) const noexcept { return cuco::pair{unset_mark(_hasher(i)), T{i}}; } @@ -104,7 +104,8 @@ template struct hash_pair_fn { CUDF_HOST_DEVICE constexpr hash_pair_fn(hash_value_type const* hashes) : _hashes{hashes} {} - __device__ __forceinline__ auto operator()(size_type i) const noexcept + __device__ __forceinline__ cuco::pair operator()( + size_type i) const noexcept { return cuco::pair{_hashes[i], IndexType{i}}; } diff --git a/cpp/src/lists/combine/concatenate_rows.cu b/cpp/src/lists/combine/concatenate_rows.cu index bef9633018bd..9a89d6cf1969 100644 --- a/cpp/src/lists/combine/concatenate_rows.cu +++ b/cpp/src/lists/combine/concatenate_rows.cu @@ -21,7 +21,6 @@ #include #include #include -#include #include namespace cudf { @@ -72,13 +71,13 @@ generate_regrouped_offsets_and_null_mask(table_device_view const& input, data_type{type_id::INT32}, input.num_rows() + 1, mask_state::UNALLOCATED, stream, mr); auto keys = - thrust::make_transform_iterator(cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [num_columns = input.num_columns()] __device__( - std::size_t i) -> size_type { return i / num_columns; })); + cuda::transform_iterator(cuda::counting_iterator{0}, + cuda::proclaim_return_type( + [num_columns = input.num_columns()] __device__( + std::size_t i) -> size_type { return i / num_columns; })); // generate sizes for the regrouped rows - auto values = thrust::make_transform_iterator( + auto values = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type([input, row_null_counts = row_null_counts.data(), @@ -153,12 +152,12 @@ rmm::device_uvector generate_null_counts(table_device_view const& inp rmm::device_uvector null_counts(input.num_rows(), stream); auto keys = - thrust::make_transform_iterator(cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [num_columns = input.num_columns()] __device__( - std::size_t i) -> size_type { return i / num_columns; })); + cuda::transform_iterator(cuda::counting_iterator{0}, + cuda::proclaim_return_type( + [num_columns = input.num_columns()] __device__( + std::size_t i) -> size_type { return i / num_columns; })); - auto null_values = thrust::make_transform_iterator( + auto null_values = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type([input] __device__(std::size_t i) -> size_type { auto const col_index = i % input.num_columns(); @@ -265,7 +264,7 @@ std::unique_ptr concatenate_rows(table_view const& input, // what we want. the data of the children will be exactly what we want, but will be grouped as if // we had concatenated all the rows together instead of concatenating within the rows. To fix // this we can simply swap in a new set of offsets that re-groups them. bmo - auto iter = thrust::make_transform_iterator( + auto iter = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [num_columns = input.num_columns(), diff --git a/cpp/src/lists/copying/scatter_helper.cu b/cpp/src/lists/copying/scatter_helper.cu index 1f95711ad186..160a7bb23a94 100644 --- a/cpp/src/lists/copying/scatter_helper.cu +++ b/cpp/src/lists/copying/scatter_helper.cu @@ -18,7 +18,6 @@ #include #include #include -#include #include namespace cudf { @@ -326,7 +325,7 @@ struct list_child_constructor { // child_list_views should now have been populated, with source and target references. - auto begin = thrust::make_transform_iterator( + auto begin = cuda::transform_iterator( child_list_views.begin(), cuda::proclaim_return_type([] __device__(auto const& row) { return row.size(); })); @@ -400,8 +399,8 @@ struct list_child_constructor { {structs_list_offsets, structs_member})); }; - auto const iter_source_member_as_list = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [&](auto child_idx) { + auto const iter_source_member_as_list = + cuda::transform_iterator(cuda::counting_iterator{0}, [&](auto child_idx) { return project_member_as_list_view(source_structs.child(child_idx), source_lists_column_view.size(), source_lists_column_view.offsets(), @@ -409,8 +408,8 @@ struct list_child_constructor { source_lists_column_view.null_count()); }); - auto const iter_target_member_as_list = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [&](auto child_idx) { + auto const iter_target_member_as_list = + cuda::transform_iterator(cuda::counting_iterator{0}, [&](auto child_idx) { return project_member_as_list_view(target_structs.child(child_idx), target_lists_column_view.size(), target_lists_column_view.offsets(), diff --git a/cpp/src/lists/dremel.cu b/cpp/src/lists/dremel.cu index dc9b17fb220f..74d5146d203c 100644 --- a/cpp/src/lists/dremel.cu +++ b/cpp/src/lists/dremel.cu @@ -291,23 +291,22 @@ dremel_data get_encoding(column_view h_col, // Merge empty at deepest parent level with the rep, def level vals at leaf level auto input_parent_rep_it = cuda::make_constant_iterator(level); - auto input_parent_def_it = - thrust::make_transform_iterator(empties_idx.begin(), - def_level_fn{d_nesting_levels + level, - d_nullability.data(), - start_at_sub_level[level], - def_at_level[level], - always_nullable}); + auto input_parent_def_it = cuda::transform_iterator(empties_idx.begin(), + def_level_fn{d_nesting_levels + level, + d_nullability.data(), + start_at_sub_level[level], + def_at_level[level], + always_nullable}); // `nesting_levels.size()` == no of list levels + leaf. Max repetition level = no of list levels auto input_child_rep_it = cuda::make_constant_iterator(nesting_levels.size() - 1); auto input_child_def_it = - thrust::make_transform_iterator(cuda::counting_iterator{column_offsets[level + 1]}, - def_level_fn{d_nesting_levels + level + 1, - d_nullability.data(), - start_at_sub_level[level + 1], - def_at_level[level + 1], - always_nullable}); + cuda::transform_iterator(cuda::counting_iterator{column_offsets[level + 1]}, + def_level_fn{d_nesting_levels + level + 1, + d_nullability.data(), + start_at_sub_level[level + 1], + def_at_level[level + 1], + always_nullable}); // Zip the input and output value iterators so that merge operation is done only once auto input_parent_zip_it = @@ -388,16 +387,15 @@ dremel_data get_encoding(column_view h_col, std::swap(temp_def_vals, def_level); // Merge empty at parent level with the rep, def level vals at current level - auto transformed_empties = thrust::make_transform_iterator(empties.begin(), offset_transformer); + auto transformed_empties = cuda::transform_iterator(empties.begin(), offset_transformer); auto input_parent_rep_it = cuda::make_constant_iterator(level); - auto input_parent_def_it = - thrust::make_transform_iterator(empties_idx.begin(), - def_level_fn{d_nesting_levels + level, - d_nullability.data(), - start_at_sub_level[level], - def_at_level[level], - always_nullable}); + auto input_parent_def_it = cuda::transform_iterator(empties_idx.begin(), + def_level_fn{d_nesting_levels + level, + d_nullability.data(), + start_at_sub_level[level], + def_at_level[level], + always_nullable}); // Zip the input and output value iterators so that merge operation is done only once auto input_parent_zip_it = diff --git a/cpp/src/lists/explode.cu b/cpp/src/lists/explode.cu index 083cfd1f3dc5..ef83718b7c6f 100644 --- a/cpp/src/lists/explode.cu +++ b/cpp/src/lists/explode.cu @@ -24,7 +24,6 @@ #include #include #include -#include #include #include @@ -50,7 +49,7 @@ std::unique_ptr
build_table( cuda::stream_ref stream, rmm::device_async_resource_ref mr) { - auto select_iter = thrust::make_transform_iterator( + auto select_iter = cuda::transform_iterator( cuda::counting_iterator{0}, [explode_column_idx](size_type i) { return i >= explode_column_idx ? i + 1 : i; }); @@ -111,7 +110,7 @@ std::unique_ptr
explode(table_view const& input_table, // Sliced columns may require rebasing of the offsets. auto offsets = explode_col.offsets_begin(); // offsets + 1 here to skip the 0th offset, which removes a - 1 operation later. - auto offsets_minus_one = thrust::make_transform_iterator( + auto offsets_minus_one = cuda::transform_iterator( cuda::std::next(offsets), cuda::proclaim_return_type([offsets] __device__(auto i) { return (i - offsets[0]) - 1; })); @@ -149,7 +148,7 @@ std::unique_ptr
explode_position(table_view const& input_table, // Sliced columns may require rebasing of the offsets. auto offsets = explode_col.offsets_begin(); // offsets + 1 here to skip the 0th offset, which removes a - 1 operation later. - auto offsets_minus_one = thrust::make_transform_iterator( + auto offsets_minus_one = cuda::transform_iterator( offsets + 1, cuda::proclaim_return_type([offsets] __device__(auto i) { return (i - offsets[0]) - 1; })); @@ -201,7 +200,7 @@ std::unique_ptr
explode_outer(table_view const& input_table, // number of nulls or empty lists found so far in the explode column rmm::device_uvector null_or_empty_offset(explode_col.size(), stream); - auto null_or_empty = thrust::make_transform_iterator( + auto null_or_empty = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [offsets, offsets_size = explode_col.size() - 1] __device__(int idx) { @@ -228,7 +227,7 @@ std::unique_ptr
explode_outer(table_view const& input_table, rmm::device_uvector pos(include_position ? gather_map_size : 0, stream, mr); // offsets + 1 here to skip the 0th offset, which removes a - 1 operation later. - auto offsets_minus_one = thrust::make_transform_iterator( + auto offsets_minus_one = cuda::transform_iterator( cuda::std::next(offsets), cuda::proclaim_return_type([offsets] __device__(auto i) { return (i - offsets[0]) - 1; })); diff --git a/cpp/src/partitioning/round_robin.cu b/cpp/src/partitioning/round_robin.cu index 6233acabd211..c9a0b26b7a0b 100644 --- a/cpp/src/partitioning/round_robin.cu +++ b/cpp/src/partitioning/round_robin.cu @@ -27,7 +27,6 @@ #include #include #include -#include #include #include @@ -78,12 +77,12 @@ std::pair, std::vector> degenerate auto nrows = input.num_rows(); // iterator for partition index rotated right by start_partition positions: - auto rotated_iter_begin = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - cuda::proclaim_return_type( - [num_partitions, start_partition] __device__(auto index) { - return (index + num_partitions - start_partition) % num_partitions; - })); + auto rotated_iter_begin = + cuda::transform_iterator(cuda::counting_iterator{0}, + cuda::proclaim_return_type( + [num_partitions, start_partition] __device__(auto index) { + return (index + num_partitions - start_partition) % num_partitions; + })); if (num_partitions == nrows) { rmm::device_uvector partition_offsets(num_partitions + 1, stream); @@ -124,7 +123,7 @@ std::pair, std::vector> degenerate // iterator for number of edges of the transposed bipartite graph; // this composes rotated_iter transform (above) iterator with // calculating number of edges of transposed bi-graph: - auto nedges_iter_begin = thrust::make_transform_iterator( + auto nedges_iter_begin = cuda::transform_iterator( rotated_iter_begin, cuda::proclaim_return_type( [nrows] __device__(auto index) { return (index < nrows ? 1 : 0); })); @@ -203,7 +202,7 @@ std::pair, std::vector> round_robin_part (start_partition - num_partitions_min_size) * max_partition_size : start_partition * (max_partition_size - 1)); - auto iter_begin = thrust::make_transform_iterator( + auto iter_begin = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type([nrows, num_partitions, @@ -239,7 +238,7 @@ std::pair, std::vector> round_robin_part // this has the effect of rotating the set of partition sizes // right by start_partition positions: // - auto rotated_iter_begin = thrust::make_transform_iterator( + auto rotated_iter_begin = cuda::transform_iterator( cuda::counting_iterator{0}, [num_partitions, start_partition, max_partition_size, num_partitions_max_size](auto index) { return ((index + num_partitions - start_partition) % num_partitions < num_partitions_max_size diff --git a/cpp/src/quantiles/quantile.cu b/cpp/src/quantiles/quantile.cu index 9bb5fe49d2b2..250bd81f41c8 100644 --- a/cpp/src/quantiles/quantile.cu +++ b/cpp/src/quantiles/quantile.cu @@ -27,7 +27,6 @@ #include #include -#include #include #include @@ -105,7 +104,7 @@ struct quantile_functor { } if (input.nullable()) { - auto sorted_validity = thrust::make_transform_iterator( + auto sorted_validity = cuda::transform_iterator( ordered_indices, cuda::proclaim_return_type( [input = *d_input] __device__(size_type idx) { return input.is_valid_nocheck(idx); })); diff --git a/cpp/src/quantiles/quantiles.cu b/cpp/src/quantiles/quantiles.cu index b8baf345df87..d3eac37020d1 100644 --- a/cpp/src/quantiles/quantiles.cu +++ b/cpp/src/quantiles/quantiles.cu @@ -21,7 +21,6 @@ #include #include -#include #include #include @@ -46,7 +45,7 @@ std::unique_ptr
quantiles(table_view const& input, auto const q_device = cudf::detail::make_device_uvector_async(q, stream, cudf::get_current_device_resource_ref()); - auto quantile_idx_iter = thrust::make_transform_iterator(q_device.begin(), quantile_idx_lookup); + auto quantile_idx_iter = cuda::transform_iterator(q_device.begin(), quantile_idx_lookup); return detail::gather(input, quantile_idx_iter, diff --git a/cpp/src/quantiles/tdigest/tdigest.cu b/cpp/src/quantiles/tdigest/tdigest.cu index cf6eaaf8654f..73fee2faccbf 100644 --- a/cpp/src/quantiles/tdigest/tdigest.cu +++ b/cpp/src/quantiles/tdigest/tdigest.cu @@ -29,7 +29,6 @@ #include #include #include -#include #include #include @@ -374,7 +373,7 @@ std::unique_ptr percentile_approx(tdigest_column_view const& input, // if any of the input digests are empty, nullify the corresponding output rows (values will be // uninitialized) auto [bitmask, null_count] = [stream, mr, &tdv]() { - auto tdigest_is_empty = thrust::make_transform_iterator( + auto tdigest_is_empty = cuda::transform_iterator( detail::size_begin(tdv), cuda::proclaim_return_type( [] __device__(size_type tdigest_size) -> size_type { return tdigest_size == 0; })); diff --git a/cpp/src/quantiles/tdigest/tdigest_aggregation.cu b/cpp/src/quantiles/tdigest/tdigest_aggregation.cu index 32c65e6f5696..a5ea03d1a8f8 100644 --- a/cpp/src/quantiles/tdigest/tdigest_aggregation.cu +++ b/cpp/src/quantiles/tdigest/tdigest_aggregation.cu @@ -37,7 +37,6 @@ #include #include #include -#include #include #include #include @@ -1027,13 +1026,13 @@ std::unique_ptr compute_tdigests(int delta, // to represent cluster indices (for example, if a tdigest had 100 clusters, the keys should fall // into the range 0-99). But since we have multiple tdigests, we need to keep the keys unique // between the groups, so we add our group start offset. - auto keys = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - compute_tdigests_keys_fn{delta, - cinfo.cluster_wl.begin(), - cinfo.cluster_start.begin(), - cinfo.num_clusters.begin(), - group_cumulative_weight}); + auto keys = + cuda::transform_iterator(cuda::counting_iterator{0}, + compute_tdigests_keys_fn{delta, + cinfo.cluster_wl.begin(), + cinfo.cluster_start.begin(), + cinfo.num_clusters.begin(), + group_cumulative_weight}); // mean and weight data auto centroid_means = cudf::make_numeric_column( @@ -1450,9 +1449,9 @@ std::unique_ptr merge_tdigests(tdigest_column_view const& tdv, auto merged_min_col = cudf::make_numeric_column( data_type{type_id::FLOAT64}, num_groups, mask_state::UNALLOCATED, stream, mr); auto min_iter = - thrust::make_transform_iterator(cuda::make_zip_iterator(cuda::std::make_tuple( - tdv.min_begin(), cudf::tdigest::detail::size_begin(tdv))), - tdigest_min{}); + cuda::transform_iterator(cuda::make_zip_iterator(cuda::std::make_tuple( + tdv.min_begin(), cudf::tdigest::detail::size_begin(tdv))), + tdigest_min{}); thrust::reduce_by_key(rmm::exec_policy_nosync(stream, cudf::get_current_device_resource_ref()), group_labels, group_labels + num_group_labels, @@ -1465,9 +1464,9 @@ std::unique_ptr merge_tdigests(tdigest_column_view const& tdv, auto merged_max_col = cudf::make_numeric_column( data_type{type_id::FLOAT64}, num_groups, mask_state::UNALLOCATED, stream, mr); auto max_iter = - thrust::make_transform_iterator(cuda::make_zip_iterator(cuda::std::make_tuple( - tdv.max_begin(), cudf::tdigest::detail::size_begin(tdv))), - tdigest_max{}); + cuda::transform_iterator(cuda::make_zip_iterator(cuda::std::make_tuple( + tdv.max_begin(), cudf::tdigest::detail::size_begin(tdv))), + tdigest_max{}); thrust::reduce_by_key(rmm::exec_policy_nosync(stream, cudf::get_current_device_resource_ref()), group_labels, group_labels + num_group_labels, diff --git a/cpp/src/reductions/all.cu b/cpp/src/reductions/all.cu index 219ac5461822..f987e34c69d6 100644 --- a/cpp/src/reductions/all.cu +++ b/cpp/src/reductions/all.cu @@ -13,7 +13,6 @@ #include #include #include -#include #include namespace cudf { @@ -54,7 +53,7 @@ struct all_fn { auto null_iter = op::min{}.template get_null_replacing_element_transformer(); auto pair_iter = cudf::dictionary::detail::make_dictionary_pair_iterator(*d_dict, input.has_nulls()); - return thrust::make_transform_iterator(pair_iter, null_iter); + return cuda::transform_iterator(pair_iter, null_iter); }(); auto d_result = cudf::detail::device_scalar(1, stream, cudf::get_current_device_resource_ref()); diff --git a/cpp/src/reductions/any.cu b/cpp/src/reductions/any.cu index 596411de969a..7682e9474b79 100644 --- a/cpp/src/reductions/any.cu +++ b/cpp/src/reductions/any.cu @@ -13,7 +13,6 @@ #include #include #include -#include #include namespace cudf { @@ -54,7 +53,7 @@ struct any_fn { auto null_iter = op::max{}.template get_null_replacing_element_transformer(); auto pair_iter = cudf::dictionary::detail::make_dictionary_pair_iterator(*d_dict, input.has_nulls()); - return thrust::make_transform_iterator(pair_iter, null_iter); + return cuda::transform_iterator(pair_iter, null_iter); }(); auto d_result = cudf::detail::device_scalar(0, stream, cudf::get_current_device_resource_ref()); diff --git a/cpp/src/reductions/compound.cuh b/cpp/src/reductions/compound.cuh index 20fd79d6aeae..c19f3055298c 100644 --- a/cpp/src/reductions/compound.cuh +++ b/cpp/src/reductions/compound.cuh @@ -14,7 +14,7 @@ #include #include -#include +#include #include #include @@ -62,19 +62,19 @@ std::unique_ptr compound_reduction(column_view const& col, if (!cudf::is_dictionary(col.type())) { if (col.has_nulls()) { - auto it = thrust::make_transform_iterator( + auto it = cuda::transform_iterator( dcol->pair_begin(), compound_op.template get_null_replacing_element_transformer()); return cudf::reduction::detail::reduce( it, col.size(), compound_op, valid_count, ddof, stream, mr); } else { - auto it = thrust::make_transform_iterator( + auto it = cuda::transform_iterator( dcol->begin(), compound_op.template get_element_transformer()); return cudf::reduction::detail::reduce( it, col.size(), compound_op, valid_count, ddof, stream, mr); } } else { - auto it = thrust::make_transform_iterator( + auto it = cuda::transform_iterator( cudf::dictionary::detail::make_dictionary_pair_iterator(*dcol, col.has_nulls()), compound_op.template get_null_replacing_element_transformer()); return cudf::reduction::detail::reduce( diff --git a/cpp/src/reductions/extrema_utils.cuh b/cpp/src/reductions/extrema_utils.cuh index ca3f945ea463..1e638a6697b5 100644 --- a/cpp/src/reductions/extrema_utils.cuh +++ b/cpp/src/reductions/extrema_utils.cuh @@ -136,7 +136,7 @@ class arg_minmax_dispatcher { Op{}.template get_null_replacing_element_transformer(); auto const p = cudf::dictionary::detail::make_dictionary_pair_iterator(*d_dict, true); - auto const it = thrust::make_transform_iterator(p, transformer); + auto const it = cuda::transform_iterator(p, transformer); return find_extremum_idx(it, input.size(), stream); } auto const it = cudf::dictionary::detail::make_dictionary_iterator(*d_dict); @@ -146,7 +146,7 @@ class arg_minmax_dispatcher { auto const d_input = column_device_view::create(input, stream); auto const transformer = Op{}.template get_null_replacing_element_transformer(); auto const it = - thrust::make_transform_iterator(d_input->pair_begin(), transformer); + cuda::transform_iterator(d_input->pair_begin(), transformer); return find_extremum_idx(it, input.size(), stream); } else { return find_extremum_idx(input.begin(), input.size(), stream); diff --git a/cpp/src/reductions/minmax.cu b/cpp/src/reductions/minmax.cu index c52af997d679..5e58e0644ee7 100644 --- a/cpp/src/reductions/minmax.cu +++ b/cpp/src/reductions/minmax.cu @@ -18,6 +18,7 @@ #include #include +#include #include #include #include diff --git a/cpp/src/reductions/nth_element.cu b/cpp/src/reductions/nth_element.cu index ea1a48ad356f..e1c2855eea7c 100644 --- a/cpp/src/reductions/nth_element.cu +++ b/cpp/src/reductions/nth_element.cu @@ -14,9 +14,9 @@ #include #include +#include #include #include -#include #include namespace cudf::reduction::detail { @@ -35,10 +35,9 @@ std::unique_ptr nth_element(column_view const& col, CUDF_EXPECTS(n >= 0 and n < valid_count, "Index out of bounds"); auto dcol = column_device_view::create(col, stream); auto bitmask_iterator = - thrust::make_transform_iterator(cudf::detail::make_validity_iterator(*dcol), - cuda::proclaim_return_type([] __device__(auto b) { - return static_cast(b); - })); + cuda::transform_iterator(cudf::detail::make_validity_iterator(*dcol), + cuda::proclaim_return_type( + [] __device__(auto b) { return static_cast(b); })); rmm::device_uvector null_skipped_index(col.size(), stream); // null skipped index for valids only. thrust::inclusive_scan(rmm::exec_policy_nosync(stream, cudf::get_current_device_resource_ref()), diff --git a/cpp/src/reductions/scan/ewm.cu b/cpp/src/reductions/scan/ewm.cu index 40a06efeccd5..171ede513fe4 100644 --- a/cpp/src/reductions/scan/ewm.cu +++ b/cpp/src/reductions/scan/ewm.cu @@ -135,7 +135,7 @@ rmm::device_uvector null_roll_up(column_view const& input, cuda rmm::device_uvector output(input.size(), stream); auto device_view = column_device_view::create(input, stream); - auto invalid_it = thrust::make_transform_iterator( + auto invalid_it = cuda::transform_iterator( cudf::detail::make_validity_iterator(*device_view), cuda::proclaim_return_type([] __device__(int valid) -> int { return 1 - valid; })); diff --git a/cpp/src/reductions/segmented/compound.cuh b/cpp/src/reductions/segmented/compound.cuh index 5923607ef287..67ab7073665f 100644 --- a/cpp/src/reductions/segmented/compound.cuh +++ b/cpp/src/reductions/segmented/compound.cuh @@ -15,8 +15,8 @@ #include #include +#include #include -#include namespace cudf { namespace reduction { @@ -66,12 +66,12 @@ std::unique_ptr compound_segmented_reduction(column_view const& col, // Run segmented reduction if (col.has_nulls()) { auto nrt = compound_op.template get_null_replacing_element_transformer(); - auto itr = thrust::make_transform_iterator(d_col->pair_begin(), nrt); + auto itr = cuda::transform_iterator(d_col->pair_begin(), nrt); cudf::reduction::detail::segmented_reduce( itr, offsets.begin(), offsets.end(), out_itr, compound_op, ddof, counts.data(), stream); } else { auto et = compound_op.template get_element_transformer(); - auto itr = thrust::make_transform_iterator(d_col->begin(), et); + auto itr = cuda::transform_iterator(d_col->begin(), et); cudf::reduction::detail::segmented_reduce( itr, offsets.begin(), offsets.end(), out_itr, compound_op, ddof, counts.data(), stream); } diff --git a/cpp/src/reductions/segmented/simple.cuh b/cpp/src/reductions/segmented/simple.cuh index 6ced8884ba94..a13a41494972 100644 --- a/cpp/src/reductions/segmented/simple.cuh +++ b/cpp/src/reductions/segmented/simple.cuh @@ -25,7 +25,6 @@ #include #include #include -#include #include #include @@ -86,12 +85,12 @@ std::unique_ptr simple_segmented_reduction( if (col.has_nulls()) { auto f = simple_op.template get_null_replacing_element_transformer(); - auto it = thrust::make_transform_iterator(dcol->pair_begin(), f); + auto it = cuda::transform_iterator(dcol->pair_begin(), f); cudf::reduction::detail::segmented_reduce( it, offsets.begin(), offsets.end(), outit, binary_op, initial_value, stream); } else { auto f = simple_op.template get_element_transformer(); - auto it = thrust::make_transform_iterator(dcol->begin(), f); + auto it = cuda::transform_iterator(dcol->begin(), f); cudf::reduction::detail::segmented_reduce( it, offsets.begin(), offsets.end(), outit, binary_op, initial_value, stream); } diff --git a/cpp/src/reductions/simple.cuh b/cpp/src/reductions/simple.cuh index c3dbb461b3df..319613e0622b 100644 --- a/cpp/src/reductions/simple.cuh +++ b/cpp/src/reductions/simple.cuh @@ -22,7 +22,6 @@ #include #include -#include #include namespace cudf { @@ -67,11 +66,11 @@ std::unique_ptr simple_reduction(column_view const& col, auto result = [&] { if (col.has_nulls()) { auto f = simple_op.template get_null_replacing_element_transformer(); - auto it = thrust::make_transform_iterator(dcol->pair_begin(), f); + auto it = cuda::transform_iterator(dcol->pair_begin(), f); return cudf::reduction::detail::reduce(it, col.size(), simple_op, initial_value, stream, mr); } else { auto f = simple_op.template get_element_transformer(); - auto it = thrust::make_transform_iterator(dcol->begin(), f); + auto it = cuda::transform_iterator(dcol->begin(), f); return cudf::reduction::detail::reduce(it, col.size(), simple_op, initial_value, stream, mr); } }(); @@ -154,7 +153,7 @@ std::unique_ptr dictionary_reduction( auto f = simple_op.template get_null_replacing_element_transformer(); auto p = cudf::dictionary::detail::make_dictionary_pair_iterator(*dcol, col.has_nulls()); - auto it = thrust::make_transform_iterator(p, f); + auto it = cuda::transform_iterator(p, f); return cudf::reduction::detail::reduce(it, col.size(), simple_op, {}, stream, mr); }(); diff --git a/cpp/src/replace/clamp.cu b/cpp/src/replace/clamp.cu index 79986de1c11e..afe115f49c27 100644 --- a/cpp/src/replace/clamp.cu +++ b/cpp/src/replace/clamp.cu @@ -34,7 +34,6 @@ #include #include #include -#include #include namespace cudf { diff --git a/cpp/src/reshape/interleave_columns.cu b/cpp/src/reshape/interleave_columns.cu index 5f80f9827697..459f4c9e8618 100644 --- a/cpp/src/reshape/interleave_columns.cu +++ b/cpp/src/reshape/interleave_columns.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -26,7 +26,6 @@ #include #include #include -#include #include namespace cudf { @@ -87,7 +86,7 @@ struct interleave_columns_impl> output_struct_members; for (size_type child_idx = 0; child_idx < num_children; ++child_idx) { // Collect children columns from the input structs columns at index `child_idx`. - auto const child_iter = thrust::make_transform_iterator( + auto const child_iter = cuda::transform_iterator( structs_columns.begin(), [&stream = stream, child_idx](auto const& col) { return structs_column_view(col).get_sliced_child(child_idx, stream); }); diff --git a/cpp/src/reshape/table_to_array.cu b/cpp/src/reshape/table_to_array.cu index 0f1a13ea89bc..9aa3fbe7a91d 100644 --- a/cpp/src/reshape/table_to_array.cu +++ b/cpp/src/reshape/table_to_array.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -23,7 +23,6 @@ #include #include #include -#include namespace cudf { namespace detail { diff --git a/cpp/src/reshape/tile.cu b/cpp/src/reshape/tile.cu index c0316ab110e6..4ce819e839f7 100644 --- a/cpp/src/reshape/tile.cu +++ b/cpp/src/reshape/tile.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -17,7 +17,7 @@ #include -#include +#include #include diff --git a/cpp/src/rolling/detail/rolling_collect_list.cuh b/cpp/src/rolling/detail/rolling_collect_list.cuh index 6eb18e42dbd5..0649e90f6811 100644 --- a/cpp/src/rolling/detail/rolling_collect_list.cuh +++ b/cpp/src/rolling/detail/rolling_collect_list.cuh @@ -18,7 +18,6 @@ #include #include #include -#include #include namespace cudf { diff --git a/cpp/src/row_operator/row_operators.cu b/cpp/src/row_operator/row_operators.cu index 698b184abef8..5bd70e191387 100644 --- a/cpp/src/row_operator/row_operators.cu +++ b/cpp/src/row_operator/row_operators.cu @@ -22,7 +22,7 @@ #include #include -#include +#include #include #include @@ -40,7 +40,7 @@ namespace { std::vector unslice_children(column_view const& c) { if (c.type().id() == type_id::STRUCT) { - auto child_it = thrust::make_transform_iterator(c.child_begin(), [](auto const& child) { + auto child_it = cuda::transform_iterator(c.child_begin(), [](auto const& child) { return column_view( child.type(), child.offset() + child.size(), // This is hacky, we don't know the actual unsliced size but diff --git a/cpp/src/sort/rank.cu b/cpp/src/sort/rank.cu index 8bf85594887a..ab626c0f02c3 100644 --- a/cpp/src/sort/rank.cu +++ b/cpp/src/sort/rank.cu @@ -26,7 +26,6 @@ #include #include #include -#include #include #include #include @@ -136,7 +135,7 @@ void tie_break_ranks_transform(cudf::device_span dense_rank_sor tie_breaker); using TransformerReturnType = cuda::std::decay_t>; - auto sorted_tied_rank = thrust::make_transform_iterator( + auto sorted_tied_rank = cuda::transform_iterator( dense_rank_sorted.begin(), cuda::proclaim_return_type( [tied_rank = tie_sorted.begin(), transformer] __device__(auto dense_pos) { diff --git a/cpp/src/strings/combine/join.cu b/cpp/src/strings/combine/join.cu index 162586ba59ab..7d0ad215743f 100644 --- a/cpp/src/strings/combine/join.cu +++ b/cpp/src/strings/combine/join.cu @@ -23,10 +23,10 @@ #include #include +#include #include #include #include -#include namespace cudf { namespace strings { @@ -158,7 +158,7 @@ std::unique_ptr join_strings(strings_column_view const& input, std::overflow_error); // build the offsets: single string output has offsets [0,chars-size] - auto sizes_itr = thrust::constant_iterator(static_cast(chars.size())); + auto sizes_itr = cuda::constant_iterator(static_cast(chars.size())); auto offsets_column = std::get<0>( cudf::strings::detail::make_offsets_child_column(sizes_itr, sizes_itr + 1, stream, mr)); diff --git a/cpp/src/strings/replace/multi.cu b/cpp/src/strings/replace/multi.cu index 29257e932029..8cb13ebfe021 100644 --- a/cpp/src/strings/replace/multi.cu +++ b/cpp/src/strings/replace/multi.cu @@ -342,9 +342,9 @@ std::unique_ptr replace_character_parallel(strings_column_view const& in auto targets_indices = rmm::device_uvector(target_count, stream); // cudf::detail::make_counting_transform_iterator hardcodes size_type - auto const copy_itr = thrust::make_transform_iterator(cuda::counting_iterator{0}, - pair_generator{fn, chars_bytes}); - auto const out_itr = cuda::make_zip_iterator( + auto const copy_itr = + cuda::transform_iterator(cuda::counting_iterator{0}, pair_generator{fn, chars_bytes}); + auto const out_itr = cuda::make_zip_iterator( cuda::std::make_tuple(targets_positions.begin(), targets_indices.begin())); auto const copy_end = cudf::detail::copy_if(copy_itr, copy_itr + chars_bytes, out_itr, copy_if_fn{}, stream); diff --git a/cpp/src/strings/search/contains_multiple.cu b/cpp/src/strings/search/contains_multiple.cu index cd8cbd2c251f..9f1d46b41049 100644 --- a/cpp/src/strings/search/contains_multiple.cu +++ b/cpp/src/strings/search/contains_multiple.cu @@ -29,7 +29,6 @@ #include #include #include -#include #include #include #include @@ -194,7 +193,7 @@ std::unique_ptr
contains_multiple(strings_column_view const& input, auto first_bytes = rmm::device_uvector(targets.size(), stream); auto indices = rmm::device_uvector(targets.size(), stream); { - auto tgt_itr = thrust::make_transform_iterator( + auto tgt_itr = cuda::transform_iterator( d_targets->begin(), cuda::proclaim_return_type([] __device__(auto const& d_tgt) -> u_char { return d_tgt.empty() ? u_char{0} : static_cast(d_tgt.data()[0]); @@ -239,7 +238,7 @@ std::unique_ptr
contains_multiple(strings_column_view const& input, auto results = std::vector>(results_iter, results_iter + targets.size()); auto d_results = [&] { auto host_results_pointer_iter = - thrust::make_transform_iterator(results.begin(), [](auto const& results_column) { + cuda::transform_iterator(results.begin(), [](auto const& results_column) { return results_column->mutable_view().template data(); }); auto host_results_pointers = diff --git a/cpp/src/strings/strings_column_factories.cu b/cpp/src/strings/strings_column_factories.cu index 7a3301a9ee0f..588b2e4c6d43 100644 --- a/cpp/src/strings/strings_column_factories.cu +++ b/cpp/src/strings/strings_column_factories.cu @@ -18,9 +18,9 @@ #include #include +#include #include #include -#include #include #include @@ -216,7 +216,7 @@ std::unique_ptr make_strings_column(device_span strin CUDF_FUNC_RANGE(); auto it_pair = - thrust::make_transform_iterator(string_views.begin(), string_view_to_pair{null_placeholder}); + cuda::transform_iterator(string_views.begin(), string_view_to_pair{null_placeholder}); return cudf::strings::detail::make_strings_column( it_pair, it_pair + string_views.size(), stream, mr); } diff --git a/cpp/src/text/deduplicate.cu b/cpp/src/text/deduplicate.cu index f8e53d0636b0..52462bcdaaf6 100644 --- a/cpp/src/text/deduplicate.cu +++ b/cpp/src/text/deduplicate.cu @@ -429,11 +429,9 @@ std::unique_ptr resolve_duplicates_pair_impl( auto const chars_span1 = cudf::device_span(d_input_chars1, chars_size1); auto const chars_span2 = cudf::device_span(d_input_chars2, chars_size2); - auto const itr1 = - thrust::make_transform_iterator(indices1.begin(), index_to_prefix_fn{chars_span1}); + auto const itr1 = cuda::transform_iterator(indices1.begin(), index_to_prefix_fn{chars_span1}); auto const end1 = itr1 + indices1.size(); - auto const itr2 = - thrust::make_transform_iterator(indices2.begin(), index_to_string_fn{chars_span2}); + auto const itr2 = cuda::transform_iterator(indices2.begin(), index_to_string_fn{chars_span2}); auto const end2 = itr2 + indices2.size(); // vectorized lower-bound and upper-bound of prefix strings improves performance of the diff --git a/cpp/src/text/normalize.cu b/cpp/src/text/normalize.cu index e52fa61b9da7..110fd12b1306 100644 --- a/cpp/src/text/normalize.cu +++ b/cpp/src/text/normalize.cu @@ -31,6 +31,7 @@ #include #include +#include #include #include #include @@ -285,7 +286,7 @@ CUDF_KERNEL void special_tokens_kernel(uint32_t* d_normalized, if (match == end) { return; } char candidate[8]; auto const ch_begin = - thrust::transform_iterator(begin, [](auto v) { return static_cast(v); }); + cuda::transform_iterator(begin, [](auto v) { return static_cast(v); }); auto const ch_end = ch_begin + cuda::std::distance(begin, match + 1); auto last = thrust::copy_if( thrust::seq, ch_begin, ch_end, candidate, [](auto c) { return c != 0 && c != ' '; }); @@ -430,7 +431,7 @@ rmm::device_uvector compute_sizes(cudf::device_span([offset] __device__(auto o) { return o - offset; })); cub::DeviceSegmentedReduce::Sum( diff --git a/cpp/src/text/vocabulary_tokenize.cu b/cpp/src/text/vocabulary_tokenize.cu index 730082f16eb4..88d8b3ab766d 100644 --- a/cpp/src/text/vocabulary_tokenize.cu +++ b/cpp/src/text/vocabulary_tokenize.cu @@ -120,7 +120,8 @@ struct tokenize_vocabulary::tokenize_vocabulary_impl { }; struct key_pair { - __device__ auto operator()(cudf::size_type idx) const noexcept + __device__ cuco::pair operator()( + cudf::size_type idx) const noexcept { return cuco::make_pair(idx, idx); } diff --git a/cpp/src/text/wordpiece_tokenize.cu b/cpp/src/text/wordpiece_tokenize.cu index 5d826d9dc55e..cf04825144bf 100644 --- a/cpp/src/text/wordpiece_tokenize.cu +++ b/cpp/src/text/wordpiece_tokenize.cu @@ -39,7 +39,6 @@ #include #include #include -#include #include namespace nvtext { @@ -172,7 +171,8 @@ namespace { * @brief Identifies the column indices as the values in the vocabulary map */ struct key_pair { - __device__ auto operator()(cudf::size_type idx) const noexcept + __device__ cuco::pair operator()( + cudf::size_type idx) const noexcept { return cuco::make_pair(idx, idx); } @@ -259,7 +259,7 @@ wordpiece_vocabulary::wordpiece_vocabulary(cudf::strings_column_view const& inpu rmm::mr::polymorphic_allocator{mr}, stream.get()); // insert them without the '##' prefix since that is how they will be looked up - auto iter_sub = thrust::make_transform_iterator(sub_map_indices.begin(), key_pair{}); + auto iter_sub = cuda::transform_iterator(sub_map_indices.begin(), key_pair{}); vocab_sub_map->insert_async(iter_sub, iter_sub + sub_map_indices.size(), stream.get()); // prefetch the [unk] vocab entry diff --git a/cpp/src/transpose/transpose.cu b/cpp/src/transpose/transpose.cu index 735137eb6d11..42cf01c244ed 100644 --- a/cpp/src/transpose/transpose.cu +++ b/cpp/src/transpose/transpose.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include @@ -18,7 +18,6 @@ #include #include -#include namespace cudf { namespace detail { @@ -40,7 +39,7 @@ std::pair, table_view> transpose(table_view const& input auto output_column = cudf::detail::interleave_columns(input, stream, mr); auto one_iter = cuda::counting_iterator{1}; - auto splits_iter = thrust::make_transform_iterator( + auto splits_iter = cuda::transform_iterator( one_iter, [width = input.num_columns()](size_type idx) { return idx * width; }); auto splits = std::vector(splits_iter, splits_iter + input.num_rows() - 1); auto output_column_views = detail::split(output_column->view(), splits, stream); diff --git a/cpp/src/utilities/linked_column.cpp b/cpp/src/utilities/linked_column.cpp index 24ce1bbca317..e751cd5189d2 100644 --- a/cpp/src/utilities/linked_column.cpp +++ b/cpp/src/utilities/linked_column.cpp @@ -1,11 +1,11 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2023, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include -#include +#include namespace cudf::detail { @@ -23,7 +23,7 @@ linked_column_view::linked_column_view(linked_column_view* parent, column_view c linked_column_view::operator column_view() const { - auto child_it = thrust::make_transform_iterator( + auto child_it = cuda::transform_iterator( children.begin(), [](auto const& c) { return static_cast(*c); }); return column_view(this->type(), this->size(), @@ -36,7 +36,7 @@ linked_column_view::operator column_view() const LinkedColVector table_to_linked_columns(table_view const& table) { - auto linked_it = thrust::make_transform_iterator( + auto linked_it = cuda::transform_iterator( table.begin(), [](auto const& c) { return std::make_shared(c); }); return LinkedColVector(linked_it, linked_it + table.num_columns()); } diff --git a/cpp/tests/column/factories_test.cpp b/cpp/tests/column/factories_test.cpp index 07c39f81e23a..3e5384a38bb4 100644 --- a/cpp/tests/column/factories_test.cpp +++ b/cpp/tests/column/factories_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/copying/copy_if_else_nested_tests.cpp b/cpp/tests/copying/copy_if_else_nested_tests.cpp index 2c383acd73b9..6dd50754dab0 100644 --- a/cpp/tests/copying/copy_if_else_nested_tests.cpp +++ b/cpp/tests/copying/copy_if_else_nested_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/copying/copy_range_tests.cpp b/cpp/tests/copying/copy_range_tests.cpp index 1f7f3dd4d5d6..ea51d4d822f6 100644 --- a/cpp/tests/copying/copy_range_tests.cpp +++ b/cpp/tests/copying/copy_range_tests.cpp @@ -15,7 +15,6 @@ #include #include -#include #include @@ -338,12 +337,12 @@ TEST_F(CopyRangeTestFixture, CopyDictionary) CUDF_TEST_EXPECT_COLUMNS_EQUAL(*decoded, expected); } - auto source_validity = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [](auto i) { return i != 3; }); - auto target_validity = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [](auto i) { return i != 3 && i != 9; }); - auto expected_validity = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [](auto i) { return i != 5 && i != 9; }); + auto source_validity = cuda::transform_iterator(cuda::counting_iterator{0}, + [](auto i) { return i != 3; }); + auto target_validity = cuda::transform_iterator(cuda::counting_iterator{0}, + [](auto i) { return i != 3 && i != 9; }); + auto expected_validity = cuda::transform_iterator(cuda::counting_iterator{0}, + [](auto i) { return i != 5 && i != 9; }); { auto source = cudf::dictionary::encode(cudf::test::strings_column_wrapper( source_elements.begin(), source_elements.end(), source_validity)); diff --git a/cpp/tests/copying/purge_nonempty_nulls_tests.cpp b/cpp/tests/copying/purge_nonempty_nulls_tests.cpp index 1e80083d7142..4650de9ee092 100644 --- a/cpp/tests/copying/purge_nonempty_nulls_tests.cpp +++ b/cpp/tests/copying/purge_nonempty_nulls_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include diff --git a/cpp/tests/copying/scatter_list_tests.cpp b/cpp/tests/copying/scatter_list_tests.cpp index 354aca593ad9..44d5a16fa084 100644 --- a/cpp/tests/copying/scatter_list_tests.cpp +++ b/cpp/tests/copying/scatter_list_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/copying/split_tests.cpp b/cpp/tests/copying/split_tests.cpp index 72faa23df0c2..f2b78b1802cf 100644 --- a/cpp/tests/copying/split_tests.cpp +++ b/cpp/tests/copying/split_tests.cpp @@ -23,7 +23,6 @@ #include #include -#include #include #include @@ -219,7 +218,7 @@ void split_custom_column(SplitFunc Split, std::vector base_strings( {"banana", "pear", "apple", "pecans", "vanilla", "cat", "mouse", "green"}); - auto string_randomizer = thrust::make_transform_iterator( + auto string_randomizer = cuda::transform_iterator( cuda::counting_iterator{0}, [&base_strings](cudf::size_type i) { return base_strings[rand() % base_strings.size()]; }); diff --git a/cpp/tests/copying/utility_tests.cpp b/cpp/tests/copying/utility_tests.cpp index 5868fc6c2096..45cf58cab6ae 100644 --- a/cpp/tests/copying/utility_tests.cpp +++ b/cpp/tests/copying/utility_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -15,7 +15,7 @@ #include #include -#include +#include #include @@ -56,9 +56,9 @@ TEST_F(EmptyLikeStringTest, ColumnStringTest) nullptr, "absent stop words"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto got = cudf::empty_like(strings); check_empty_string_columns(got->view(), strings); diff --git a/cpp/tests/dictionary/factories_test.cpp b/cpp/tests/dictionary/factories_test.cpp index 876bd3c74de4..4aa5db548b56 100644 --- a/cpp/tests/dictionary/factories_test.cpp +++ b/cpp/tests/dictionary/factories_test.cpp @@ -11,7 +11,7 @@ #include #include -#include +#include struct DictionaryFactoriesTest : public cudf::test::BaseFixture {}; @@ -32,7 +32,7 @@ TEST_F(DictionaryFactoriesTest, ColumnViewsWithNulls) cudf::test::fixed_width_column_wrapper keys{-11.75, 4.25, 7.125, 0.5, 12.0}; std::vector h_values{1, 3, 2, 0, 1, 4, 1}; cudf::test::fixed_width_column_wrapper indices( - h_values.begin(), h_values.end(), thrust::make_transform_iterator(h_values.begin(), [](auto v) { + h_values.begin(), h_values.end(), cuda::transform_iterator(h_values.begin(), [](auto v) { return v > 0; })); auto dictionary = cudf::make_dictionary_column(keys, indices); diff --git a/cpp/tests/dictionary/remove_keys_test.cpp b/cpp/tests/dictionary/remove_keys_test.cpp index 412fcd419588..3af990abf959 100644 --- a/cpp/tests/dictionary/remove_keys_test.cpp +++ b/cpp/tests/dictionary/remove_keys_test.cpp @@ -13,7 +13,7 @@ #include #include -#include +#include #include @@ -35,7 +35,7 @@ TEST_F(DictionaryRemoveKeysTest, StringsColumn) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto const decoded = cudf::dictionary::decode(result->view()); CUDF_TEST_EXPECT_COLUMNS_EQUAL(decoded->view(), expected); } diff --git a/cpp/tests/dictionary/set_keys_test.cpp b/cpp/tests/dictionary/set_keys_test.cpp index 2e178c9fadac..510153202224 100644 --- a/cpp/tests/dictionary/set_keys_test.cpp +++ b/cpp/tests/dictionary/set_keys_test.cpp @@ -12,7 +12,7 @@ #include #include -#include +#include #include @@ -34,7 +34,7 @@ TEST_F(DictionarySetKeysTest, StringsKeys) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto decoded = cudf::dictionary::decode(result->view()); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*decoded, expected); } diff --git a/cpp/tests/groupby/rank_scan_tests.cpp b/cpp/tests/groupby/rank_scan_tests.cpp index a123093fd528..fc673ec312e2 100644 --- a/cpp/tests/groupby/rank_scan_tests.cpp +++ b/cpp/tests/groupby/rank_scan_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/io/csv_test.cpp b/cpp/tests/io/csv_test.cpp index fd56b7a2e7b0..ee7d8dcfe8be 100644 --- a/cpp/tests/io/csv_test.cpp +++ b/cpp/tests/io/csv_test.cpp @@ -362,7 +362,7 @@ TYPED_TEST(CsvFixedPointWriterTest, SingleColumnNegativeScale) reference_strings.end(), cuda::counting_iterator{0}, std::back_inserter(valid_reference_strings), - validity.functor()); + [](std::size_t i) { return (i % 2) == 0; }); reference_strings = valid_reference_strings; using DecimalType = TypeParam; @@ -408,7 +408,7 @@ TYPED_TEST(CsvFixedPointWriterTest, SingleColumnPositiveScale) reference_strings.end(), cuda::counting_iterator{0}, std::back_inserter(valid_reference_strings), - validity.functor()); + [](std::size_t i) { return (i % 2) == 0; }); reference_strings = valid_reference_strings; using DecimalType = TypeParam; diff --git a/cpp/tests/io/experimental/hybrid_scan_common.cpp b/cpp/tests/io/experimental/hybrid_scan_common.cpp index 1bb4c0374b49..cf372e845b5b 100644 --- a/cpp/tests/io/experimental/hybrid_scan_common.cpp +++ b/cpp/tests/io/experimental/hybrid_scan_common.cpp @@ -41,8 +41,8 @@ cudf::test::strings_column_wrapper constant_strings(cudf::size_type value) { CUDF_EXPECTS(value >= 0 && value <= 9999, "String value must be between 0000 and 9999"); - auto elements = thrust::make_transform_iterator(cuda::make_constant_iterator(value), - [](auto i) { return std::format("{:04d}", i); }); + auto elements = cuda::transform_iterator(cuda::make_constant_iterator(value), + [](auto i) { return std::format("{:04d}", i); }); return cudf::test::strings_column_wrapper(elements, elements + num_ordered_rows); } diff --git a/cpp/tests/io/orc_chunked_reader_test.cu b/cpp/tests/io/orc_chunked_reader_test.cu index 8fb00b629ae9..ee8c6c634870 100644 --- a/cpp/tests/io/orc_chunked_reader_test.cu +++ b/cpp/tests/io/orc_chunked_reader_test.cu @@ -36,6 +36,8 @@ #include #include +#include + namespace { enum class output_limit : std::size_t {}; enum class input_limit : std::size_t {}; @@ -60,8 +62,10 @@ auto write_file(std::vector>& input_columns, { if (nullable) { // Generate deterministic bitmask instead of random bitmask for easy computation of data size. - auto const valid_iter = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) { return i % 4 != 3; }); + auto const valid_values = + std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) -> bool { return i % 4 != 3; }); + auto const valid_iter = valid_values.begin(); cudf::size_type offset{0}; for (auto& col : input_columns) { auto const [null_mask, null_count] = @@ -439,12 +443,14 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithString) // 20000 rows of 1 char each (20000 + 80004) = B0 100004 100004 // 20000 rows of 4 chars each (80000 + 80004) = B1 160004 260008 // 20000 rows of 16 chars each (320000 + 80004) = B2 400004 660012 - auto const strings = std::vector{"a", "bbbb", "cccccccccccccccc"}; - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - if (i < 20000) { return strings[0]; } - if (i < 40000) { return strings[1]; } - return strings[2]; - }); + auto const strings = std::vector{"a", "bbbb", "cccccccccccccccc"}; + auto const str_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) -> std::string { + if (i < 20000) { return strings[0]; } + if (i < 40000) { return strings[1]; } + return strings[2]; + }); + auto const str_iter = str_values.begin(); input_columns.emplace_back(strings_col(str_iter, str_iter + num_rows).release()); // Cumulative sizes: @@ -538,9 +544,11 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithStructs) auto child1 = int32s_col(int_iter, int_iter + num_rows); auto child2 = int32s_col(int_iter + num_rows, int_iter + num_rows * 2); - auto const str_iter = cudf::detail::make_counting_transform_iterator( - 0, [&](int32_t i) { return std::to_string(i); }); - auto child3 = strings_col{str_iter, str_iter + num_rows}; + auto const str_values = + std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) -> std::string { return std::to_string(i); }); + auto const str_iter = str_values.begin(); + auto child3 = strings_col{str_iter, str_iter + num_rows}; return structs_col{{child1, child2, child3}}.release(); }()); @@ -625,9 +633,10 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithListsNoNulls) auto const template_lists = int32s_lists_col{ int32s_lists_col{}, int32s_lists_col{0}, int32s_lists_col{1, 2}, int32s_lists_col{3, 4, 5}}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) -> int32_t { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); input_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -711,9 +720,10 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithListsHavingNulls) int32s_lists_col{0}, int32s_lists_col{1, 2}, int32s_lists_col{3, 4, 5, 6, 7, 8, 9} /* this list will be nullified out */}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) -> int32_t { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); input_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -791,16 +801,19 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithStructsOfLists) child_columns.emplace_back( int32s_col(int_iter + num_rows, int_iter + num_rows * 2).release()); - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - return std::to_string(i) + "++++++++++++++++++++" + std::to_string(i); - }); + auto const str_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) -> std::string { + return std::to_string(i) + "++++++++++++++++++++" + std::to_string(i); + }); + auto const str_iter = str_values.begin(); child_columns.emplace_back(strings_col{str_iter, str_iter + num_rows}.release()); auto const template_lists = int32s_lists_col{ int32s_lists_col{}, int32s_lists_col{0}, int32s_lists_col{0, 1}, int32s_lists_col{0, 1, 2}}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) -> int32_t { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); child_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -925,9 +938,12 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadWithListsOfStructs) auto child1 = int32s_col(int_iter, int_iter + num_structs); auto child2 = int32s_col(int_iter + num_structs, int_iter + num_structs * 2); - auto const str_iter = cudf::detail::make_counting_transform_iterator( - 0, [&](int32_t i) { return std::to_string(i) + std::to_string(i) + std::to_string(i); }); - auto child3 = strings_col{str_iter, str_iter + num_structs}; + auto const str_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) -> std::string { + return std::to_string(i) + std::to_string(i) + std::to_string(i); + }); + auto const str_iter = str_values.begin(); + auto child3 = strings_col{str_iter, str_iter + num_structs}; return structs_col{{child1, child2, child3}}.release(); }; @@ -1036,9 +1052,10 @@ TEST_F(OrcChunkedReaderTest, TestChunkedReadNullCount) { auto constexpr num_rows = 100'000; - auto const sequence = cuda::constant_iterator(1); - auto const validity = - cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 4 != 3; }); + auto const sequence = cuda::constant_iterator(1); + auto const validity_values = std::views::iota(cudf::size_type{0}) | + std::views::transform([](auto i) -> bool { return i % 4 != 3; }); + auto const validity = validity_values.begin(); std::vector> cols; cols.push_back(int32s_col{sequence, sequence + num_rows, validity}.release()); auto const expected = std::make_unique(std::move(cols)); @@ -1163,16 +1180,20 @@ TEST_F(OrcChunkedReaderInputLimitTest, MixedColumns) auto const iter1 = cuda::counting_iterator{0}; auto const col1 = int32s_col(iter1, iter1 + num_rows); - auto const iter2 = - cudf::detail::make_counting_transform_iterator(0, [](int i) { return static_cast(i); }); - auto const col2 = doubles_col(iter2, iter2 + num_rows); - - auto const strings = std::vector{"abc", "de", "fghi"}; - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - if (i < 250000) { return strings[0]; } - if (i < 750000) { return strings[1]; } - return strings[2]; - }); + auto const values2 = std::views::iota(0) | std::views::transform([](int i) -> double { + return static_cast(i); + }); + auto const iter2 = values2.begin(); + auto const col2 = doubles_col(iter2, iter2 + num_rows); + + auto const strings = std::vector{"abc", "de", "fghi"}; + auto const str_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) -> std::string { + if (i < 250000) { return strings[0]; } + if (i < 750000) { return strings[1]; } + return strings[2]; + }); + auto const str_iter = str_values.begin(); auto const col3 = strings_col(str_iter, str_iter + num_rows); auto const filename = std::string{"mixed_columns"}; @@ -1444,11 +1465,12 @@ TEST_F(OrcChunkedReaderInputLimitTest, SizeTypeRowsOverflow) int64_t constexpr total_rows = num_rows * num_reps; static_assert(total_rows > std::numeric_limits::max()); - auto const it = - thrust::make_transform_iterator(cuda::counting_iterator{0}, [num_rows](int64_t i) { + auto const values = + std::views::iota(int64_t{0}, num_rows) | + std::views::transform([num_rows](int64_t i) -> data_type { return (i % num_rows) % static_cast(std::numeric_limits::max() / 2); }); - auto const col = data_col(it, it + num_rows); + auto const col = data_col(values.begin(), values.end()); auto const chunk_table = cudf::table_view{{col}}; std::vector data_buffer; @@ -1480,8 +1502,9 @@ TEST_F(OrcChunkedReaderInputLimitTest, SizeTypeRowsOverflow) // Check validity of the last 5 million rows. auto const sequence_start = num_rows_to_skip % num_rows; - auto const skipped_col = data_col(it + sequence_start, it + sequence_start + num_rows_to_read); - auto const expected = cudf::table_view{{skipped_col}}; + auto const skipped_col = + data_col(values.begin() + sequence_start, values.begin() + sequence_start + num_rows_to_read); + auto const expected = cudf::table_view{{skipped_col}}; auto const read_opts = cudf::io::orc_reader_options::builder( @@ -1578,7 +1601,9 @@ TEST_P(OrcChunkedDecompressionTest, RoundTripBasic) auto const num_rows = 12'345; std::vector> input_columns; - auto value_iter = cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i / 4; }); + auto const values = + std::views::iota(int32_t{0}) | std::views::transform([](auto i) -> int32_t { return i / 4; }); + auto const value_iter = values.begin(); input_columns.emplace_back(int32s_col(value_iter, value_iter + num_rows).release()); input_columns.emplace_back(int64s_col(value_iter, value_iter + num_rows).release()); auto expected = std::make_unique(std::move(input_columns)); diff --git a/cpp/tests/io/parquet_chunked_reader_test.cu b/cpp/tests/io/parquet_chunked_reader_test.cu index 5096920b97e7..a95217572813 100644 --- a/cpp/tests/io/parquet_chunked_reader_test.cu +++ b/cpp/tests/io/parquet_chunked_reader_test.cu @@ -38,8 +38,10 @@ #include #include +#include #include #include +#include #include namespace { @@ -60,8 +62,9 @@ auto write_file(std::vector>& input_columns, { if (nullable) { // Generate deterministic bitmask instead of random bitmask for easy computation of data size. - auto const valid_iter = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) { return i % 4 != 3; }); + auto const valid_values = std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) { return i % 4 != 3; }); + auto const valid_iter = valid_values.begin(); cudf::size_type offset{0}; for (auto& col : input_columns) { @@ -355,12 +358,13 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithString) // 20000 rows of 1 char each (20000 + 80004) = B0 100004 100004 // 20000 rows of 4 chars each (80000 + 80004) = B1 160004 260008 // 20000 rows of 16 chars each (320000 + 80004) = B2 400004 660012 - auto const strings = std::vector{"a", "bbbb", "cccccccccccccccc"}; - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - if (i < 20000) { return strings[0]; } - if (i < 40000) { return strings[1]; } - return strings[2]; - }); + auto const strings = std::vector{"a", "bbbb", "cccccccccccccccc"}; + auto const str_values = std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { + if (i < 20000) { return strings[0]; } + if (i < 40000) { return strings[1]; } + return strings[2]; + }); + auto const str_iter = str_values.begin(); input_columns.emplace_back(strings_col(str_iter, str_iter + num_rows).release()); // Cumulative sizes: @@ -610,9 +614,10 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithStringPrecise) // 20000 rows alternating 1-4 chars each (50000 + 80004) A0 130004 130004 // 20000 rows alternating 1-4 chars each (50000 + 80004) A1 130004 260008 // ... - auto const strings = std::vector{"a", "bbbb"}; - auto const str_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return strings[i % 2]; }); + auto const strings = std::vector{"a", "bbbb"}; + auto const str_values = std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) { return strings[i % 2]; }); + auto const str_iter = str_values.begin(); input_columns.emplace_back(strings_col(str_iter, str_iter + num_rows).release()); // Cumulative sizes: @@ -659,9 +664,10 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithStructs) auto child1 = int32s_col(int_iter, int_iter + num_rows); auto child2 = int32s_col(int_iter + num_rows, int_iter + num_rows * 2); - auto const str_iter = cudf::detail::make_counting_transform_iterator( - 0, [&](int32_t i) { return std::to_string(i); }); - auto child3 = strings_col{str_iter, str_iter + num_rows}; + auto const str_values = std::views::iota(int32_t{0}) | + std::views::transform([&](int32_t i) { return std::to_string(i); }); + auto const str_iter = str_values.begin(); + auto child3 = strings_col{str_iter, str_iter + num_rows}; return structs_col{{child1, child2, child3}}.release(); }()); @@ -742,9 +748,10 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithListsNoNulls) auto const template_lists = int32s_lists_col{ int32s_lists_col{}, int32s_lists_col{0}, int32s_lists_col{1, 2}, int32s_lists_col{3, 4, 5}}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); input_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -832,9 +839,10 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithListsHavingNulls) int32s_lists_col{0}, int32s_lists_col{1, 2}, int32s_lists_col{3, 4, 5, 6, 7, 8, 9} /* this list will be nullified out */}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); input_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -921,16 +929,19 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithStructsOfLists) child_columns.emplace_back( int32s_col(int_iter + num_rows, int_iter + num_rows * 2).release()); - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - return std::to_string(i) + "++++++++++++++++++++" + std::to_string(i); - }); + auto const str_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { + return std::to_string(i) + "++++++++++++++++++++" + std::to_string(i); + }); + auto const str_iter = str_values.begin(); child_columns.emplace_back(strings_col{str_iter, str_iter + num_rows}.release()); auto const template_lists = int32s_lists_col{ int32s_lists_col{}, int32s_lists_col{0}, int32s_lists_col{0, 1}, int32s_lists_col{0, 1, 2}}; - auto const gather_iter = - cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { return i % 4; }); - auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); + auto const gather_values = + std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { return i % 4; }); + auto const gather_iter = gather_values.begin(); + auto const gather_map = int32s_col(gather_iter, gather_iter + num_rows); child_columns.emplace_back( std::move(cudf::gather(cudf::table_view{{template_lists}}, gather_map)->release().front())); @@ -1064,9 +1075,11 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadWithListsOfStructs) auto child1 = int32s_col(int_iter, int_iter + num_structs); auto child2 = int32s_col(int_iter + num_structs, int_iter + num_structs * 2); - auto const str_iter = cudf::detail::make_counting_transform_iterator( - 0, [&](int32_t i) { return std::to_string(i) + std::to_string(i) + std::to_string(i); }); - auto child3 = strings_col{str_iter, str_iter + num_structs}; + auto const str_values = std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { + return std::to_string(i) + std::to_string(i) + std::to_string(i); + }); + auto const str_iter = str_values.begin(); + auto child3 = strings_col{str_iter, str_iter + num_structs}; return structs_col{{child1, child2, child3}}.release(); }; @@ -1184,8 +1197,9 @@ TEST_F(ParquetChunkedReaderTest, TestChunkedReadNullCount) auto constexpr num_rows = 100'000; auto const sequence = cuda::constant_iterator{1}; - auto const validity = - cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 4 != 3; }); + auto const validity_values = + std::views::iota(cudf::size_type{0}) | std::views::transform([](auto i) { return i % 4 != 3; }); + auto const validity = validity_values.begin(); cudf::test::fixed_width_column_wrapper col{sequence, sequence + num_rows, validity}; std::vector> cols; cols.push_back(col.release()); @@ -1305,17 +1319,19 @@ TEST_F(ParquetChunkedReaderInputLimitConstrainedTest, MixedColumns) auto iter1 = cuda::counting_iterator{0}; cudf::test::fixed_width_column_wrapper col1(iter1, iter1 + num_rows); - auto iter2 = - cudf::detail::make_counting_transform_iterator(0, [](int i) { return static_cast(i); }); + auto const values2 = + std::views::iota(0) | std::views::transform([](int i) { return static_cast(i); }); + auto const iter2 = values2.begin(); cudf::test::fixed_width_column_wrapper col2(iter2, iter2 + num_rows); - auto const strings = std::vector{"abc", "de", "fghi"}; - auto const str_iter = cudf::detail::make_counting_transform_iterator(0, [&](int32_t i) { - if (i < 250000) { return strings[0]; } - if (i < 750000) { return strings[1]; } - return strings[2]; - }); - auto col3 = strings_col(str_iter, str_iter + num_rows); + auto const strings = std::vector{"abc", "de", "fghi"}; + auto const str_values = std::views::iota(int32_t{0}) | std::views::transform([&](int32_t i) { + if (i < 250000) { return strings[0]; } + if (i < 750000) { return strings[1]; } + return strings[2]; + }); + auto const str_iter = str_values.begin(); + auto col3 = strings_col(str_iter, str_iter + num_rows); auto tbl = cudf::table_view{{col1, col2, col3}}; @@ -1985,8 +2001,9 @@ TEST_F(ParquetChunkedReaderTest, TestNumRowsPerSourceMultipleSources) std::equal(expected_counts.cbegin(), expected_counts.cend(), num_rows_per_source.cbegin())); // Expected source index column - auto const src_index = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) { return i / num_rows; }); + auto const src_indices = std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) { return i / num_rows; }); + auto const src_index = src_indices.begin(); auto const expected_src_index = cudf::test::fixed_width_column_wrapper( src_index, src_index + nsources * num_rows); CUDF_TEST_EXPECT_COLUMNS_EQUAL(result->view().column(0), expected_src_index); @@ -2153,12 +2170,15 @@ TEST_F(ParquetChunkedReaderTest, TestRowIndexColumnMultipleSources) EXPECT_EQ(result->num_columns(), 3); // Expected source and (file-local) row index columns - auto const src_index = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) { return i / num_rows; }); + auto const src_indices = std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) { return i / num_rows; }); + auto const src_index = src_indices.begin(); auto const expected_src_index = cudf::test::fixed_width_column_wrapper( src_index, src_index + nsources * num_rows); - auto const row_index = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) -> size_t { return i % num_rows; }); + auto const row_indices = + std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) -> size_t { return i % num_rows; }); + auto const row_index = row_indices.begin(); auto const expected_row_index = cudf::test::fixed_width_column_wrapper(row_index, row_index + nsources * num_rows); CUDF_TEST_EXPECT_COLUMNS_EQUAL(result->view().column(0), expected_src_index); @@ -2190,12 +2210,15 @@ TEST_F(ParquetChunkedReaderTest, TestRowIndexColumnMultipleSources) EXPECT_EQ(result->num_rows(), rows_to_read); // Expected source and (file-local) row index columns - auto const src_index = cudf::detail::make_counting_transform_iterator( - rows_to_skip, [](cudf::size_type i) { return i / num_rows; }); + auto const src_indices = std::views::iota(static_cast(rows_to_skip)) | + std::views::transform([](cudf::size_type i) { return i / num_rows; }); + auto const src_index = src_indices.begin(); auto const expected_src_index = cudf::test::fixed_width_column_wrapper(src_index, src_index + rows_to_read); - auto const row_index = cudf::detail::make_counting_transform_iterator( - rows_to_skip, [](cudf::size_type i) -> size_t { return i % num_rows; }); + auto const row_indices = + std::views::iota(static_cast(rows_to_skip)) | + std::views::transform([](cudf::size_type i) -> size_t { return i % num_rows; }); + auto const row_index = row_indices.begin(); auto const expected_row_index = cudf::test::fixed_width_column_wrapper(row_index, row_index + rows_to_read); CUDF_TEST_EXPECT_COLUMNS_EQUAL(result->view().column(0), expected_src_index); @@ -2305,19 +2328,21 @@ TEST_F(ParquetReaderTest, BooleanList) // Validity helpers std::bernoulli_distribution bn(0.7f); - auto valids = - cudf::detail::make_counting_transform_iterator(0, [&](int index) { return bn(gen); }); - auto list_valids = - cudf::detail::make_counting_transform_iterator(0, [&](int index) { return index % 100; }); + auto valids = std::vector(num_rows); + std::generate(valids.begin(), valids.end(), [&] { return bn(gen); }); + auto const list_valid_values = + std::views::iota(0) | std::views::transform([](int index) { return index % 100; }); + auto const list_valids = list_valid_values.begin(); // str(non-nullable) auto col0 = testdata::ascending(); // list(nullable) - auto bools_iter = - cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 2; }); - auto bools_col = - cudf::test::fixed_width_column_wrapper(bools_iter, bools_iter + num_rows, valids); + auto const bool_values = + std::views::iota(0) | std::views::transform([](auto i) { return i % 2; }); + auto const bools_iter = bool_values.begin(); + auto bools_col = cudf::test::fixed_width_column_wrapper( + bools_iter, bools_iter + num_rows, valids.begin()); auto offsets_iter = cuda::counting_iterator{0}; auto offsets_col = cudf::test::fixed_width_column_wrapper( offsets_iter, offsets_iter + num_rows + 1); @@ -2534,8 +2559,10 @@ TEST_F(ParquetChunkedReaderTest, ReadStringsWithRowBounds) // Generate strings of varying length (~6-10 chars each) to produce enough // data for the chunked reader to return more than one chunk. - auto const str_iter = cudf::detail::make_counting_transform_iterator( - 0, [](cudf::size_type i) { return "str_" + std::to_string(i); }); + auto const str_values = + std::views::iota(cudf::size_type{0}) | + std::views::transform([](cudf::size_type i) { return "str_" + std::to_string(i); }); + auto const str_iter = str_values.begin(); std::vector> input_columns; input_columns.emplace_back(strings_col(str_iter, str_iter + num_rows).release()); diff --git a/cpp/tests/iterator/optional_iterator_test_numeric.cu b/cpp/tests/iterator/optional_iterator_test_numeric.cu index 22f8f064f110..62a2f9eb361a 100644 --- a/cpp/tests/iterator/optional_iterator_test_numeric.cu +++ b/cpp/tests/iterator/optional_iterator_test_numeric.cu @@ -8,9 +8,9 @@ #include +#include #include #include -#include #include #include @@ -93,7 +93,7 @@ TYPED_TEST(NumericOptionalIteratorTest, mean_var_output) // GPU test auto it_dev = d_col->template optional_begin(cudf::nullate::YES{}); - auto it_dev_squared = thrust::make_transform_iterator(it_dev, transformer); + auto it_dev_squared = cuda::transform_iterator(it_dev, transformer); // this can be computed with a single reduce and without a temporary output vector // but the approach increases the compile time by ~2x diff --git a/cpp/tests/iterator/pair_iterator_test_numeric.cu b/cpp/tests/iterator/pair_iterator_test_numeric.cu index ac536f37693a..b73868ec5075 100644 --- a/cpp/tests/iterator/pair_iterator_test_numeric.cu +++ b/cpp/tests/iterator/pair_iterator_test_numeric.cu @@ -8,8 +8,8 @@ #include +#include #include -#include #include using TestingTypes = cudf::test::NumericTypes; @@ -104,7 +104,7 @@ TYPED_TEST(NumericPairIteratorTest, mean_var_output) // GPU test auto it_dev = d_col->template pair_begin(); - auto it_dev_squared = thrust::make_transform_iterator(it_dev, transformer); + auto it_dev_squared = cuda::transform_iterator(it_dev, transformer); auto result = thrust::reduce(rmm::exec_policy_nosync(cudf::get_default_stream()), it_dev_squared, it_dev_squared + d_col->size(), diff --git a/cpp/tests/iterator/value_iterator_test_transform.cu b/cpp/tests/iterator/value_iterator_test_transform.cu index 8551d04e4535..3fce5be970fd 100644 --- a/cpp/tests/iterator/value_iterator_test_transform.cu +++ b/cpp/tests/iterator/value_iterator_test_transform.cu @@ -7,8 +7,8 @@ #include #include +#include #include -#include struct TransformedIteratorTest : public IteratorTest {}; @@ -56,7 +56,7 @@ TEST_F(TransformedIteratorTest, null_iterator_upcast) // GPU test auto it_dev = cudf::detail::make_null_replacement_iterator(*d_col, T{0}); - auto it_dev_upcast = thrust::make_transform_iterator(it_dev, cast_fn{}); + auto it_dev_upcast = cuda::transform_iterator(it_dev, cast_fn{}); this->iterator_test_thrust(replaced_array, it_dev_upcast, d_col->size()); this->iterator_test_cub(expected_value, it_dev, d_col->size()); } @@ -99,8 +99,8 @@ TEST_F(TransformedIteratorTest, null_iterator_square) // GPU test auto it_dev = cudf::detail::make_null_replacement_iterator(*d_col, T{0}); - auto it_dev_upcast = thrust::make_transform_iterator(it_dev, cast_fn{}); - auto it_dev_squared = thrust::make_transform_iterator(it_dev_upcast, transformer); + auto it_dev_upcast = cuda::transform_iterator(it_dev, cast_fn{}); + auto it_dev_squared = cuda::transform_iterator(it_dev_upcast, transformer); this->iterator_test_thrust(replaced_array, it_dev_squared, d_col->size()); this->iterator_test_cub(expected_value, it_dev_squared, d_col->size()); } diff --git a/cpp/tests/join/semi_anti_join_tests.cpp b/cpp/tests/join/semi_anti_join_tests.cpp index 11fb4af2a35c..dd1786036eb1 100644 --- a/cpp/tests/join/semi_anti_join_tests.cpp +++ b/cpp/tests/join/semi_anti_join_tests.cpp @@ -24,8 +24,8 @@ #include #include +#include #include -#include #include #include diff --git a/cpp/tests/lists/combine/concatenate_list_elements_tests.cpp b/cpp/tests/lists/combine/concatenate_list_elements_tests.cpp index 849cb79f4932..59ff3c1c39ea 100644 --- a/cpp/tests/lists/combine/concatenate_list_elements_tests.cpp +++ b/cpp/tests/lists/combine/concatenate_list_elements_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/lists/contains_tests.cpp b/cpp/tests/lists/contains_tests.cpp index e83da89df49b..1f7ea0f79092 100644 --- a/cpp/tests/lists/contains_tests.cpp +++ b/cpp/tests/lists/contains_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2025, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 * */ diff --git a/cpp/tests/lists/stream_compaction/apply_boolean_mask_tests.cpp b/cpp/tests/lists/stream_compaction/apply_boolean_mask_tests.cpp index 4a782ad8e913..36be64551bb5 100644 --- a/cpp/tests/lists/stream_compaction/apply_boolean_mask_tests.cpp +++ b/cpp/tests/lists/stream_compaction/apply_boolean_mask_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include diff --git a/cpp/tests/partitioning/hash_partition_test.cpp b/cpp/tests/partitioning/hash_partition_test.cpp index 0a9d4f4992fc..df3ae6fcbbec 100644 --- a/cpp/tests/partitioning/hash_partition_test.cpp +++ b/cpp/tests/partitioning/hash_partition_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include @@ -22,7 +22,6 @@ #include #include -#include #include @@ -239,11 +238,10 @@ TEST_F(HashPartition, LargePartitionCountWithNulls) cudf::size_type const num_rows = 200; cudf::size_type const num_partitions = partitions_exceeding_shared_memory(); auto const iter = cuda::counting_iterator{0}; - auto const valids = thrust::make_transform_iterator(iter, [](auto i) { return i % 4 != 0; }); + auto const valids = cuda::transform_iterator(iter, [](auto i) { return i % 4 != 0; }); fixed_width_column_wrapper ints(iter, iter + num_rows, valids); - auto const str_iter = - thrust::make_transform_iterator(iter, [](auto i) { return std::to_string(i); }); + auto const str_iter = cuda::transform_iterator(iter, [](auto i) { return std::to_string(i); }); strings_column_wrapper strings(str_iter, str_iter + num_rows, valids); auto const input = cudf::table_view({ints, strings}); @@ -396,7 +394,7 @@ void run_fixed_width_test(size_t cols, if (has_nulls) { std::generate_n(std::back_inserter(columns), cols, [rows]() { auto iter = cuda::counting_iterator{0}; - auto valids = thrust::make_transform_iterator(iter, [](auto i) { return i % 4 != 0; }); + auto valids = cuda::transform_iterator(iter, [](auto i) { return i % 4 != 0; }); return fixed_width_column_wrapper(iter, iter + rows, valids); }); } else { diff --git a/cpp/tests/reductions/distinct_count_tests.cpp b/cpp/tests/reductions/distinct_count_tests.cpp index fd4ce0c9725c..12655805adaf 100644 --- a/cpp/tests/reductions/distinct_count_tests.cpp +++ b/cpp/tests/reductions/distinct_count_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/reductions/rank_tests.cpp b/cpp/tests/reductions/rank_tests.cpp index 5e197a25ca54..de64083a6b20 100644 --- a/cpp/tests/reductions/rank_tests.cpp +++ b/cpp/tests/reductions/rank_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/reductions/unique_count_tests.cpp b/cpp/tests/reductions/unique_count_tests.cpp index fcf2bfce8011..bb364acfbeee 100644 --- a/cpp/tests/reductions/unique_count_tests.cpp +++ b/cpp/tests/reductions/unique_count_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/reshape/byte_cast_tests.cpp b/cpp/tests/reshape/byte_cast_tests.cpp index cbf6874ea52f..a6b0e4105768 100644 --- a/cpp/tests/reshape/byte_cast_tests.cpp +++ b/cpp/tests/reshape/byte_cast_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/rolling/collect_ops_test.cpp b/cpp/tests/rolling/collect_ops_test.cpp index 224edcd28647..c0710f051549 100644 --- a/cpp/tests/rolling/collect_ops_test.cpp +++ b/cpp/tests/rolling/collect_ops_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/rolling/grouped_rolling_range_test.cpp b/cpp/tests/rolling/grouped_rolling_range_test.cpp index 13a10083048d..f190cd453614 100644 --- a/cpp/tests/rolling/grouped_rolling_range_test.cpp +++ b/cpp/tests/rolling/grouped_rolling_range_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -403,9 +403,9 @@ struct GroupedRollingRangeOrderByDecimalTypedTest /// For scale == 2, the rep values are: [0, 1, 2, 3, ... 11, 12, 13] [[nodiscard]] column_ptr generate_order_by_column(numeric::scale_type scale) const { - auto const begin = thrust::make_transform_iterator( - cuda::counting_iterator{0}, - [&](auto i) -> Rep { return (i * 10000) / base::pow10[scale + 2]; }); + auto const begin = + cuda::transform_iterator(cuda::counting_iterator{0}, + [&](auto i) -> Rep { return (i * 10000) / base::pow10[scale + 2]; }); return decimals_column{begin, begin + num_rows, numeric::scale_type{scale}}.release(); } diff --git a/cpp/tests/rolling/lead_lag_test.cpp b/cpp/tests/rolling/lead_lag_test.cpp index 5a4072cf02de..126b3ef4679d 100644 --- a/cpp/tests/rolling/lead_lag_test.cpp +++ b/cpp/tests/rolling/lead_lag_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include @@ -17,7 +17,7 @@ #include #include -#include +#include #include #include diff --git a/cpp/tests/rolling/nth_element_test.cpp b/cpp/tests/rolling/nth_element_test.cpp index 51856100dc17..eda384ff18c5 100644 --- a/cpp/tests/rolling/nth_element_test.cpp +++ b/cpp/tests/rolling/nth_element_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,7 +12,7 @@ #include #include -#include +#include #include #include diff --git a/cpp/tests/rolling/range_rolling_window_test.cpp b/cpp/tests/rolling/range_rolling_window_test.cpp index 35aa8beb0ce8..e73291c1114c 100644 --- a/cpp/tests/rolling/range_rolling_window_test.cpp +++ b/cpp/tests/rolling/range_rolling_window_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,6 @@ #include #include -#include #include @@ -91,7 +90,7 @@ void verify_results_for_ascending(WindowExecT exec) auto const n_rows = exec.num_rows(); auto const all_valid = cuda::make_constant_iterator(true); auto const all_invalid = cuda::make_constant_iterator(false); - auto const last_invalid = thrust::make_transform_iterator( + auto const last_invalid = cuda::transform_iterator( cuda::counting_iterator{0}, [&n_rows](auto i) { return i != (n_rows - 1); }); CUDF_TEST_EXPECT_COLUMNS_EQUAL( @@ -174,8 +173,8 @@ void verify_results_for_descending(WindowExecT exec) { auto const all_valid = cuda::make_constant_iterator(true); auto const all_invalid = cuda::make_constant_iterator(false); - auto const first_invalid = thrust::make_transform_iterator( - cuda::counting_iterator{0}, [](auto i) { return i != 0; }); + auto const first_invalid = cuda::transform_iterator(cuda::counting_iterator{0}, + [](auto i) { return i != 0; }); CUDF_TEST_EXPECT_COLUMNS_EQUAL( exec(cudf::make_count_aggregation(cudf::null_policy::INCLUDE)) diff --git a/cpp/tests/search/search_test.cpp b/cpp/tests/search/search_test.cpp index 520f07c88b7f..f7bfefad5e42 100644 --- a/cpp/tests/search/search_test.cpp +++ b/cpp/tests/search/search_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,8 +14,8 @@ #include +#include #include -#include #include #include @@ -647,14 +647,12 @@ TEST_F(SearchTest, empty_table_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 0, 0, 0, 0, 0, 0, 0, 0, 0}; @@ -676,14 +674,12 @@ TEST_F(SearchTest, empty_values_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{}; @@ -706,14 +702,12 @@ TEST_F(SearchTest, non_null_column__find_first_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 0, 0, 1, 2, 3, 3, 4, 4, 5}; @@ -736,14 +730,12 @@ TEST_F(SearchTest, non_null_column__find_last_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 0, 1, 1, 3, 3, 4, 4, 5, 5}; @@ -766,14 +758,12 @@ TEST_F(SearchTest, non_null_column_desc__find_first_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{5, 5, 4, 4, 2, 2, 1, 1, 0, 0}; @@ -796,14 +786,12 @@ TEST_F(SearchTest, non_null_column_desc__find_last_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{5, 5, 5, 4, 3, 2, 2, 1, 1, 0}; @@ -826,14 +814,12 @@ TEST_F(SearchTest, nullable_column__find_last__nulls_as_smallest_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{2, 2, 3, 3, 5, 5, 6, 6, 7, 7}; @@ -856,14 +842,12 @@ TEST_F(SearchTest, nullable_column__find_first__nulls_as_smallest_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 2, 2, 3, 4, 5, 5, 6, 6, 7}; @@ -886,14 +870,12 @@ TEST_F(SearchTest, nullable_column__find_last__nulls_as_largest_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 1, 1, 3, 3, 4, 4, 5, 5, 7}; @@ -935,14 +917,12 @@ TEST_F(SearchTest, nullable_column__find_first__nulls_as_largest_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values( h_val_strings.begin(), h_val_strings.end(), - thrust::make_transform_iterator(h_val_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_strings.begin(), [](auto str) { return str != nullptr; })); fixed_width_column_wrapper expect{0, 0, 1, 2, 3, 3, 4, 4, 5, 5}; @@ -978,26 +958,22 @@ TEST_F(SearchTest, table__find_first_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1047,26 +1023,22 @@ TEST_F(SearchTest, table__find_last_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1116,26 +1088,22 @@ TEST_F(SearchTest, table_partial_desc__find_first_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1186,26 +1154,22 @@ TEST_F(SearchTest, table_partial_desc__find_last_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1253,26 +1217,22 @@ TEST_F(SearchTest, table__find_first__nulls_as_smallest_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1320,26 +1280,22 @@ TEST_F(SearchTest, table__find_last__nulls_as_smallest_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1387,26 +1343,22 @@ TEST_F(SearchTest, table__find_first__nulls_as_largest_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1454,26 +1406,22 @@ TEST_F(SearchTest, table__find_last__nulls_as_largest_string) cudf::test::strings_column_wrapper column_0( h_col_0_strings.begin(), h_col_0_strings.end(), - thrust::make_transform_iterator(h_col_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper column_2( h_col_2_strings.begin(), h_col_2_strings.end(), - thrust::make_transform_iterator(h_col_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_2_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_0( h_val_0_strings.begin(), h_val_0_strings.end(), - thrust::make_transform_iterator(h_val_0_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_0_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper values_2( h_val_2_strings.begin(), h_val_2_strings.end(), - thrust::make_transform_iterator(h_val_2_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_val_2_strings.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(column_0.release()); @@ -1509,8 +1457,7 @@ TEST_F(SearchTest, contains_true_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool expect = true; bool result = false; @@ -1528,8 +1475,7 @@ TEST_F(SearchTest, contains_false_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool expect = false; bool result = false; @@ -1547,8 +1493,7 @@ TEST_F(SearchTest, contains_empty_value_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool expect = false; bool result = false; @@ -1566,8 +1511,7 @@ TEST_F(SearchTest, contains_empty_column_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool expect = false; bool result = false; @@ -1585,8 +1529,7 @@ TEST_F(SearchTest, contains_nullable_column_true_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool result = false; bool expect = true; @@ -1604,8 +1547,7 @@ TEST_F(SearchTest, contains_nullable_column_false_string) cudf::test::strings_column_wrapper column( h_col_strings.begin(), h_col_strings.end(), - thrust::make_transform_iterator(h_col_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_col_strings.begin(), [](auto str) { return str != nullptr; })); bool result = false; bool expect = false; @@ -1715,14 +1657,12 @@ TEST_F(SearchTest, multi_contains_some_string_with_nulls) cudf::test::strings_column_wrapper haystack( h_haystack_strings.begin(), h_haystack_strings.end(), - thrust::make_transform_iterator(h_haystack_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_haystack_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper needles( h_needles_strings.begin(), h_needles_strings.end(), - thrust::make_transform_iterator(h_needles_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_needles_strings.begin(), [](auto str) { return str != nullptr; })); auto result = cudf::contains(haystack, needles); @@ -1739,14 +1679,12 @@ TEST_F(SearchTest, multi_contains_none_string_with_nulls) cudf::test::strings_column_wrapper haystack( h_haystack_strings.begin(), h_haystack_strings.end(), - thrust::make_transform_iterator(h_haystack_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_haystack_strings.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper needles( h_needles_strings.begin(), h_needles_strings.end(), - thrust::make_transform_iterator(h_needles_strings.begin(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_needles_strings.begin(), [](auto str) { return str != nullptr; })); auto result = cudf::contains(haystack, needles); diff --git a/cpp/tests/stream_compaction/distinct_tests.cpp b/cpp/tests/stream_compaction/distinct_tests.cpp index 8fac53ca64b1..32253437be8c 100644 --- a/cpp/tests/stream_compaction/distinct_tests.cpp +++ b/cpp/tests/stream_compaction/distinct_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/stream_compaction/stable_distinct_tests.cpp b/cpp/tests/stream_compaction/stable_distinct_tests.cpp index f5e5bbe2a13a..120aa12dabac 100644 --- a/cpp/tests/stream_compaction/stable_distinct_tests.cpp +++ b/cpp/tests/stream_compaction/stable_distinct_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2023-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2023-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/stream_compaction/unique_tests.cpp b/cpp/tests/stream_compaction/unique_tests.cpp index cf73f4e739e2..2bc0480e47e4 100644 --- a/cpp/tests/stream_compaction/unique_tests.cpp +++ b/cpp/tests/stream_compaction/unique_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/streams/copying_test.cpp b/cpp/tests/streams/copying_test.cpp index 502f34606a9f..80ac01aca25f 100644 --- a/cpp/tests/streams/copying_test.cpp +++ b/cpp/tests/streams/copying_test.cpp @@ -340,7 +340,7 @@ TEST_F(CopyingTest, ContiguousSplit) std::vector base_strings( {"banana", "pear", "apple", "pecans", "vanilla", "cat", "mouse", "green"}); - auto string_randomizer = thrust::make_transform_iterator( + auto string_randomizer = cuda::transform_iterator( cuda::counting_iterator{0}, [&base_strings](cudf::size_type i) { return base_strings[rand() % base_strings.size()]; }); diff --git a/cpp/tests/streams/strings/attributes_test.cpp b/cpp/tests/streams/strings/attributes_test.cpp index d3c1765642c2..af8026fbbc95 100644 --- a/cpp/tests/streams/strings/attributes_test.cpp +++ b/cpp/tests/streams/strings/attributes_test.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,15 +10,17 @@ #include #include +#include + struct StringsAttributesTest : public cudf::test::BaseFixture {}; TEST_F(StringsAttributesTest, CodePoints) { std::vector h_strings{"eee", "bb", nullptr, "", "aa", "bbb", "ééé"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::code_points(strings_view, cudf::test::get_default_stream()); @@ -39,9 +41,9 @@ TEST_F(StringsAttributesTest, CountBytes) std::vector h_strings{ "eee", "bb", nullptr, "", "aa", "ééé", "something a bit longer than 32 bytes"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::count_bytes(strings_view, cudf::test::get_default_stream()); diff --git a/cpp/tests/strings/array_tests.cpp b/cpp/tests/strings/array_tests.cpp index 5ccd725df8b2..0f9d16329f0c 100644 --- a/cpp/tests/strings/array_tests.cpp +++ b/cpp/tests/strings/array_tests.cpp @@ -16,7 +16,6 @@ #include #include -#include #include @@ -61,8 +60,7 @@ TEST_P(SliceParmsTest, Slice) cudf::test::strings_column_wrapper expected( h_strings.begin() + start, h_strings.begin() + end, - thrust::make_transform_iterator(h_strings.begin() + start, - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin() + start, [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(*results, expected); } @@ -81,8 +79,7 @@ TEST_P(SliceParmsTest, SliceAllNulls) cudf::test::strings_column_wrapper expected( h_strings.begin() + start, h_strings.begin() + end, - thrust::make_transform_iterator(h_strings.begin() + start, - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin() + start, [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/attrs_tests.cpp b/cpp/tests/strings/attrs_tests.cpp index ecf6bfe93728..01939440c816 100644 --- a/cpp/tests/strings/attrs_tests.cpp +++ b/cpp/tests/strings/attrs_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,7 +11,7 @@ #include #include -#include +#include #include @@ -21,9 +21,9 @@ TEST_F(StringsAttributesTest, CodePoints) { std::vector h_strings{"eee", "bb", nullptr, "", "aa", "bbb", "ééé"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); { @@ -55,9 +55,9 @@ TEST_F(StringsAttributesTest, StringsLengths) std::vector h_strings{ "eee", "bb", nullptr, "", "aa", "ééé", "something a bit longer than 32 bytes"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); { @@ -66,7 +66,7 @@ TEST_F(StringsAttributesTest, StringsLengths) cudf::test::fixed_width_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -75,7 +75,7 @@ TEST_F(StringsAttributesTest, StringsLengths) cudf::test::fixed_width_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/booleans_tests.cpp b/cpp/tests/strings/booleans_tests.cpp index 0675502128ff..9377a031430e 100644 --- a/cpp/tests/strings/booleans_tests.cpp +++ b/cpp/tests/strings/booleans_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2023, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,7 +10,7 @@ #include #include -#include +#include #include @@ -20,9 +20,9 @@ TEST_F(StringsConvertTest, ToBooleans) { std::vector h_strings{"false", nullptr, "", "true", "True", "False"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto true_scalar = cudf::string_scalar("true"); @@ -30,9 +30,9 @@ TEST_F(StringsConvertTest, ToBooleans) std::vector h_expected{false, false, false, true, false, false}; cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -40,15 +40,15 @@ TEST_F(StringsConvertTest, FromBooleans) { std::vector h_strings{"true", nullptr, "false", "true", "true", "false"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector h_column{true, false, false, true, true, false}; cudf::test::fixed_width_column_wrapper column( - h_column.begin(), - h_column.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_column.begin(), h_column.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto true_scalar = cudf::string_scalar("true"); auto false_scalar = cudf::string_scalar("false"); diff --git a/cpp/tests/strings/case_tests.cpp b/cpp/tests/strings/case_tests.cpp index e264468c689f..2d3c35c1f149 100644 --- a/cpp/tests/strings/case_tests.cpp +++ b/cpp/tests/strings/case_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,7 +12,7 @@ #include #include -#include +#include #include @@ -26,9 +26,9 @@ TEST_F(StringsCaseTest, ToLower) "éxamples abc", "123 456", nullptr, "are the", "tést strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::to_lower(strings_view); @@ -36,7 +36,7 @@ TEST_F(StringsCaseTest, ToLower) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -48,9 +48,9 @@ TEST_F(StringsCaseTest, ToUpper) "ÉXAMPLES ABC", "123 456", nullptr, "ARE THE", "TÉST STRINGS", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::to_upper(strings_view); @@ -58,7 +58,7 @@ TEST_F(StringsCaseTest, ToUpper) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -70,9 +70,9 @@ TEST_F(StringsCaseTest, Swapcase) "éXAMPLES AbC", "123 456", nullptr, "are the", "TÉST STRINGS", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::swapcase(strings_view); @@ -80,7 +80,7 @@ TEST_F(StringsCaseTest, Swapcase) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/chars_types_tests.cpp b/cpp/tests/strings/chars_types_tests.cpp index c61b88564fab..7f99f1168d9a 100644 --- a/cpp/tests/strings/chars_types_tests.cpp +++ b/cpp/tests/strings/chars_types_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,7 +11,7 @@ #include #include -#include +#include #include #include @@ -58,9 +58,9 @@ TEST_P(CharsTypes, AllTypes) auto is_parm = GetParam(); cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::all_characters_of_type(strings_view, is_parm); @@ -75,7 +75,7 @@ TEST_P(CharsTypes, AllTypes) cudf::test::fixed_width_column_wrapper expected( sub_expected, sub_expected + strings_count, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -130,9 +130,9 @@ TEST_F(StringsCharsTest, Alphanumeric) "\t\r\n\f "}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::all_characters_of_type( @@ -155,9 +155,9 @@ TEST_F(StringsCharsTest, Alphanumeric) true, false}; cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -181,9 +181,9 @@ TEST_F(StringsCharsTest, AlphaNumericSpace) "\t\r\n\f "}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto types = @@ -208,9 +208,9 @@ TEST_F(StringsCharsTest, AlphaNumericSpace) true, true}; cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -234,9 +234,9 @@ TEST_F(StringsCharsTest, Numerics) "\t\r\n\f "}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto types = cudf::strings::string_character_types::DIGIT | @@ -262,9 +262,9 @@ TEST_F(StringsCharsTest, Numerics) false, false}; cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/combine/concatenate_tests.cpp b/cpp/tests/strings/combine/concatenate_tests.cpp index 3e081c7638fa..a84208364fa7 100644 --- a/cpp/tests/strings/combine/concatenate_tests.cpp +++ b/cpp/tests/strings/combine/concatenate_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include #include -#include +#include constexpr cudf::test::debug_output_level verbosity{cudf::test::debug_output_level::ALL_ERRORS}; @@ -25,12 +25,12 @@ TEST_F(StringsCombineTest, Concatenate) cudf::test::strings_column_wrapper strings1( h_strings1.begin(), h_strings1.end(), - thrust::make_transform_iterator(h_strings1.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings1.begin(), [](auto str) { return str != nullptr; })); std::vector h_strings2{"xyz", "abc", "d", "éa", "", nullptr, "f"}; cudf::test::strings_column_wrapper strings2( h_strings2.begin(), h_strings2.end(), - thrust::make_transform_iterator(h_strings2.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings2.begin(), [](auto str) { return str != nullptr; })); std::vector strings_columns; strings_columns.push_back(strings1); @@ -43,7 +43,7 @@ TEST_F(StringsCombineTest, Concatenate) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::concatenate(table); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -54,7 +54,7 @@ TEST_F(StringsCombineTest, Concatenate) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::concatenate(table, cudf::string_scalar(":")); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -64,7 +64,7 @@ TEST_F(StringsCombineTest, Concatenate) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::concatenate(table, cudf::string_scalar(":"), cudf::string_scalar("_")); @@ -75,7 +75,7 @@ TEST_F(StringsCombineTest, Concatenate) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::concatenate(table, cudf::string_scalar(""), cudf::string_scalar("")); diff --git a/cpp/tests/strings/combine/join_strings_tests.cpp b/cpp/tests/strings/combine/join_strings_tests.cpp index 4b99c2e4aa2c..0c781d49c91c 100644 --- a/cpp/tests/strings/combine/join_strings_tests.cpp +++ b/cpp/tests/strings/combine/join_strings_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include #include -#include +#include struct JoinStringsTest : public cudf::test::BaseFixture {}; @@ -21,9 +21,9 @@ TEST_F(JoinStringsTest, Join) { std::vector h_strings{"eee", "bb", nullptr, "zzzz", "", "aaa", "ééé"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto view1 = cudf::strings_column_view(strings); { diff --git a/cpp/tests/strings/contains_tests.cpp b/cpp/tests/strings/contains_tests.cpp index bbc285e85add..53e5d9f18829 100644 --- a/cpp/tests/strings/contains_tests.cpp +++ b/cpp/tests/strings/contains_tests.cpp @@ -18,7 +18,6 @@ #include #include -#include #include #include @@ -65,9 +64,9 @@ TEST_F(StringsContainsTests, ContainsTest) ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); std::vector patterns{"\\d", @@ -141,7 +140,7 @@ TEST_F(StringsContainsTests, ContainsTest) cudf::test::fixed_width_column_wrapper expected( h_expected, h_expected + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); auto prog = cudf::strings::regex_program::create(ptn); auto results = cudf::strings::contains_re(strings_view, *prog); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -153,16 +152,16 @@ TEST_F(StringsContainsTests, MatchesTest) std::vector h_strings{ "The quick brown @fox jumps", "ovér the", "lazy @dog", "1234", "00:0:00", nullptr, ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); { auto const pattern = std::string("lazy"); cudf::test::fixed_width_column_wrapper expected( {false, false, true, false, false, false, false}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); auto prog = cudf::strings::regex_program::create(pattern); auto results = cudf::strings::matches_re(strings_view, *prog); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -171,7 +170,7 @@ TEST_F(StringsContainsTests, MatchesTest) auto const pattern = std::string("\\d+"); cudf::test::fixed_width_column_wrapper expected( {false, false, false, true, true, false, false}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); auto prog = cudf::strings::regex_program::create(pattern); auto results = cudf::strings::matches_re(strings_view, *prog); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -180,7 +179,7 @@ TEST_F(StringsContainsTests, MatchesTest) auto const pattern = std::string("@\\w+"); cudf::test::fixed_width_column_wrapper expected( {false, false, false, false, false, false, false}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); auto prog = cudf::strings::regex_program::create(pattern); auto results = cudf::strings::matches_re(strings_view, *prog); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); @@ -189,7 +188,7 @@ TEST_F(StringsContainsTests, MatchesTest) auto const pattern = std::string(".*"); cudf::test::fixed_width_column_wrapper expected( {true, true, true, true, true, false, true}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); auto prog = cudf::strings::regex_program::create(pattern); auto results = cudf::strings::matches_re(strings_view, *prog); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); diff --git a/cpp/tests/strings/datetime_tests.cpp b/cpp/tests/strings/datetime_tests.cpp index 57c635c5180d..b6251d1651e5 100644 --- a/cpp/tests/strings/datetime_tests.cpp +++ b/cpp/tests/strings/datetime_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include #include -#include +#include #include @@ -30,9 +30,9 @@ TEST_F(StringsDatetimeTest, ToTimestamp) "1921-01-07T14:32:07Z", "1969-12-31T23:59:45Z"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector h_expected{ 131246625, 1563399277, 0, 0, 1553085296, 1582934400, -1545730073, -15}; @@ -41,9 +41,9 @@ TEST_F(StringsDatetimeTest, ToTimestamp) strings_view, cudf::data_type{cudf::type_id::TIMESTAMP_SECONDS}, "%Y-%m-%dT%H:%M:%SZ"); cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); results = cudf::strings::is_timestamp(strings_view, "%Y-%m-%dT%H:%M:%SZ"); @@ -331,14 +331,14 @@ TEST_F(StringsDatetimeTest, FromTimestamp) cudf::test::fixed_width_column_wrapper timestamps( h_timestamps.begin(), h_timestamps.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::from_timestamps(timestamps); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/durations_tests.cpp b/cpp/tests/strings/durations_tests.cpp index acad5be9ed37..292ea50c9158 100644 --- a/cpp/tests/strings/durations_tests.cpp +++ b/cpp/tests/strings/durations_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,7 +11,7 @@ #include #include -#include +#include #include #include @@ -34,14 +34,14 @@ TEST_F(StringsDurationsTest, FromToDurations) cudf::test::fixed_width_column_wrapper durations( h_durations.begin(), h_durations.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::from_durations(durations); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); // @@ -394,16 +394,16 @@ TEST_F(StringsDurationsTest, ParseSingle) ""}; // error auto size = cudf::column_view(string_src).size(); std::array expected_v{0, 0, 1, -1, 23, -23, 59, -59, 99, -99, 0, 1, 0}; - auto it1 = thrust::make_transform_iterator(expected_v.data(), - [](auto i) { return cudf::duration_s{i * 3600}; }); + auto it1 = + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i * 3600}; }); cudf::test::fixed_width_column_wrapper expected_s1(it1, it1 + size); auto results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), "%H"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s1); - auto it2 = thrust::make_transform_iterator(expected_v.data(), - [](auto i) { return cudf::duration_s{i * 60}; }); + auto it2 = + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i * 60}; }); cudf::test::fixed_width_column_wrapper expected_s2(it2, it2 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -411,15 +411,15 @@ TEST_F(StringsDurationsTest, ParseSingle) CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s2); auto it3 = - thrust::make_transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); cudf::test::fixed_width_column_wrapper expected_s3(it3, it3 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), "%S"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s3); - auto it4 = thrust::make_transform_iterator(expected_v.data(), - [](auto i) { return cudf::duration_ms{i * 60000}; }); + auto it4 = cuda::transform_iterator(expected_v.data(), + [](auto i) { return cudf::duration_ms{i * 60000}; }); cudf::test::fixed_width_column_wrapper expected_ms(it4, it4 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -458,15 +458,15 @@ TEST_F(StringsDurationsTest, ParseMultiple) 3661, 0}; auto it1 = - thrust::make_transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); cudf::test::fixed_width_column_wrapper expected_s1(it1, it1 + size); auto results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), "%H:%M:%S"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s1); - auto it2 = thrust::make_transform_iterator( - expected_v.data(), [](auto i) { return cudf::duration_D{i / (24 * 3600)}; }); + auto it2 = cuda::transform_iterator(expected_v.data(), + [](auto i) { return cudf::duration_D{i / (24 * 3600)}; }); cudf::test::fixed_width_column_wrapper expected_D2(it2, it2 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -512,15 +512,15 @@ TEST_F(StringsDurationsTest, ParseSubsecond) (3661) * 1000000000L, 0}; auto it1 = - thrust::make_transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_ns{i}; }); + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_ns{i}; }); cudf::test::fixed_width_column_wrapper expected_ns1(it1, it1 + size); auto results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), "%H:%M:%S"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_ns1); - auto it2 = thrust::make_transform_iterator(expected_v.data(), - [](auto i) { return cudf::duration_ms{i / 1000000}; }); + auto it2 = cuda::transform_iterator(expected_v.data(), + [](auto i) { return cudf::duration_ms{i / 1000000}; }); cudf::test::fixed_width_column_wrapper expected_ms2(it2, it2 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -567,15 +567,15 @@ TEST_F(StringsDurationsTest, ParseAMPM) 0, 0}; auto it1 = - thrust::make_transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); cudf::test::fixed_width_column_wrapper expected_s1(it1, it1 + size); auto results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), "%H:%M:%S %p"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s1); - auto it2 = thrust::make_transform_iterator( - expected_v.data(), [](auto i) { return cudf::duration_D{i / (24 * 3600)}; }); + auto it2 = cuda::transform_iterator(expected_v.data(), + [](auto i) { return cudf::duration_D{i / (24 * 3600)}; }); cudf::test::fixed_width_column_wrapper expected_D2(it2, it2 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -619,7 +619,7 @@ TEST_F(StringsDurationsTest, ParseCompoundSpecifier) 0, 0}; auto it1 = - thrust::make_transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_s{i}; }); cudf::test::fixed_width_column_wrapper expected_s1(it1, it1 + size); auto results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), @@ -631,8 +631,8 @@ TEST_F(StringsDurationsTest, ParseCompoundSpecifier) "%OI:%OM:%OS %p"); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_s1); - auto it2 = thrust::make_transform_iterator(expected_v.data(), - [](auto i) { return cudf::duration_ms{i * 1000}; }); + auto it2 = + cuda::transform_iterator(expected_v.data(), [](auto i) { return cudf::duration_ms{i * 1000}; }); cudf::test::fixed_width_column_wrapper expected_s2(it2, it2 + size); results = cudf::strings::to_durations(cudf::strings_column_view(string_src), cudf::data_type(cudf::type_to_id()), diff --git a/cpp/tests/strings/extract_tests.cpp b/cpp/tests/strings/extract_tests.cpp index 8766fa3534a7..460e73c0e05d 100644 --- a/cpp/tests/strings/extract_tests.cpp +++ b/cpp/tests/strings/extract_tests.cpp @@ -15,7 +15,7 @@ #include #include -#include +#include #include @@ -27,9 +27,9 @@ TEST_F(StringsExtractTests, ExtractTest) "First Last", "Joe Schmoe", "John Smith", "Jane Smith", "Beyonce", "Sting", nullptr, ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); std::vector h_expecteds{"First", @@ -54,12 +54,12 @@ TEST_F(StringsExtractTests, ExtractTest) cudf::test::strings_column_wrapper expected1( h_expecteds.data(), h_expecteds.data() + h_strings.size(), - thrust::make_transform_iterator(h_expecteds.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expecteds.begin(), [](auto str) { return str != nullptr; })); cudf::test::strings_column_wrapper expected2( h_expecteds.data() + h_strings.size(), h_expecteds.data() + h_expecteds.size(), - thrust::make_transform_iterator(h_expecteds.data() + h_strings.size(), - [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expecteds.data() + h_strings.size(), + [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(expected1.release()); columns.push_back(expected2.release()); @@ -246,9 +246,9 @@ TEST_F(StringsExtractTests, EmptyExtractTest) { std::vector h_strings{nullptr, "AAA", "AAA_A", "AAA_AAA_", "A__", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto pattern = std::string("([^_]*)\\Z"); @@ -257,7 +257,7 @@ TEST_F(StringsExtractTests, EmptyExtractTest) cudf::test::strings_column_wrapper expected( h_expected.data(), h_expected.data() + h_strings.size(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); std::vector> columns; columns.push_back(expected.release()); cudf::table table_expected(std::move(columns)); @@ -298,7 +298,7 @@ TEST_F(StringsExtractTests, ExtractAllTest) std::vector h_input( {"123 banana 7 eleven", "41 apple", "6 péar 0 pair", nullptr, "", "bees", "4 paré"}); auto validity = - thrust::make_transform_iterator(h_input.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_input.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper input(h_input.begin(), h_input.end(), validity); auto sv = cudf::strings_column_view(input); @@ -388,9 +388,9 @@ TEST_F(StringsExtractTests, MediumRegex) "abcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnop" "qrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyz"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::extract(strings_view, *prog); @@ -398,7 +398,7 @@ TEST_F(StringsExtractTests, MediumRegex) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(results->get_column(0), expected); } @@ -418,9 +418,9 @@ TEST_F(StringsExtractTests, LargeRegex) "abcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnop" "qrstuvwxyzabcdefghijklmnopqrstuvwxyzabcdefghijklmnopqrstuvwxyz"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::extract(strings_view, *prog); @@ -428,7 +428,7 @@ TEST_F(StringsExtractTests, LargeRegex) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(results->get_column(0), expected); } @@ -472,6 +472,6 @@ TEST_F(StringsExtractTests, CrlfLineAnchorExtNewline) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(results->get_column(0), expected); } diff --git a/cpp/tests/strings/find_multiple_tests.cpp b/cpp/tests/strings/find_multiple_tests.cpp index 3932a53feb4f..2a128ebd142b 100644 --- a/cpp/tests/strings/find_multiple_tests.cpp +++ b/cpp/tests/strings/find_multiple_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include #include -#include +#include #include @@ -23,9 +23,9 @@ TEST_F(StringsFindMultipleTest, FindMultiple) { std::vector h_strings{"Héllo", "thesé", nullptr, "lease", "test strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); std::vector h_targets{"é", "a", "e", "i", "o", "u", "es"}; diff --git a/cpp/tests/strings/find_tests.cpp b/cpp/tests/strings/find_tests.cpp index 09358a9f7966..105d1caf6267 100644 --- a/cpp/tests/strings/find_tests.cpp +++ b/cpp/tests/strings/find_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,6 +14,8 @@ #include #include +#include + #include struct StringsFindTest : public cudf::test::BaseFixture {}; @@ -258,9 +260,9 @@ TEST_F(StringsFindTest, StartsWith) { std::vector h_targets{"éa", "th", "e", "ll", nullptr, ""}; cudf::test::strings_column_wrapper targets( - h_targets.begin(), - h_targets.end(), - thrust::make_transform_iterator(h_targets.begin(), [](auto str) { return str != nullptr; })); + h_targets.begin(), h_targets.end(), cuda::transform_iterator(h_targets.begin(), [](auto str) { + return str != nullptr; + })); auto targets_view = cudf::strings_column_view(targets); cudf::test::fixed_width_column_wrapper expected({0, 1, 0, 0, 0, 1}, @@ -300,9 +302,9 @@ TEST_F(StringsFindTest, EndsWith) { std::vector h_targets{"éa", "sé", "th", nullptr, "tést strings", ""}; cudf::test::strings_column_wrapper targets( - h_targets.begin(), - h_targets.end(), - thrust::make_transform_iterator(h_targets.begin(), [](auto str) { return str != nullptr; })); + h_targets.begin(), h_targets.end(), cuda::transform_iterator(h_targets.begin(), [](auto str) { + return str != nullptr; + })); auto targets_view = cudf::strings_column_view(targets); cudf::test::fixed_width_column_wrapper expected({0, 1, 0, 0, 1, 1}, diff --git a/cpp/tests/strings/findall_tests.cpp b/cpp/tests/strings/findall_tests.cpp index e4999a41bcc7..357b237f3ab0 100644 --- a/cpp/tests/strings/findall_tests.cpp +++ b/cpp/tests/strings/findall_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,7 +14,7 @@ #include #include -#include +#include struct StringsFindallTests : public cudf::test::BaseFixture {}; diff --git a/cpp/tests/strings/floats_tests.cpp b/cpp/tests/strings/floats_tests.cpp index 059217cb259b..721d49ed08f1 100644 --- a/cpp/tests/strings/floats_tests.cpp +++ b/cpp/tests/strings/floats_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,7 +11,7 @@ #include #include -#include +#include #include @@ -61,9 +61,9 @@ TEST_F(StringsConvertTest, ToFloats32) "123abc", "456e", "-1.78e+5", "-122.33644782123456789", "12e+309", "3.4028236E38", "INF", "Infinity"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector h_expected; std::for_each(h_strings.begin(), h_strings.end(), [&](char const* str) { @@ -74,9 +74,9 @@ TEST_F(StringsConvertTest, ToFloats32) auto results = cudf::strings::to_floats(strings_view, cudf::data_type{cudf::type_id::FLOAT32}); cudf::test::fixed_width_column_wrapper expected( - h_expected.begin(), - h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_expected.begin(), h_expected.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(*results, expected); } @@ -95,16 +95,16 @@ TEST_F(StringsConvertTest, FromFloats32) "100.0", "654321.25", "-12761.125", "0.0", "5.0", "-4.0", "NaN", "8.395422433e+11", "-0.0"}; cudf::test::fixed_width_column_wrapper floats( - h_floats.begin(), - h_floats.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + h_floats.begin(), h_floats.end(), cuda::transform_iterator(h_expected.begin(), [](auto str) { + return str != nullptr; + })); auto results = cudf::strings::from_floats(floats); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(*results, expected); } @@ -158,16 +158,16 @@ TEST_F(StringsConvertTest, FromFloats64) "100.0", "654321.25", "-12761.125", "0.0", "5.0", "-4.0", "NaN", "8.395422232e+11", "-0.0"}; cudf::test::fixed_width_column_wrapper floats( - h_floats.begin(), - h_floats.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + h_floats.begin(), h_floats.end(), cuda::transform_iterator(h_expected.begin(), [](auto str) { + return str != nullptr; + })); auto results = cudf::strings::from_floats(floats); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(*results, expected); } diff --git a/cpp/tests/strings/integers_tests.cpp b/cpp/tests/strings/integers_tests.cpp index 75838579d555..46a90bd880c8 100644 --- a/cpp/tests/strings/integers_tests.cpp +++ b/cpp/tests/strings/integers_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -15,8 +15,8 @@ #include #include +#include #include -#include #include #include @@ -76,15 +76,15 @@ TEST_F(StringsConvertTest, IsIntegerBoundCheckWithNulls) std::vector const h_strings{ "eee", "1234", nullptr, "", "-9832", "93.24", "765é", nullptr}; auto const strings = cudf::test::strings_column_wrapper( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto const results = cudf::strings::is_integer(cudf::strings_column_view(strings), cudf::data_type{cudf::type_id::INT32}); // Input has null elements then the output should have the same null mask auto const expected = cudf::test::fixed_width_column_wrapper( std::initializer_list{0, 1, NULL_VAL, 0, 1, 0, 0, NULL_VAL}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -189,15 +189,15 @@ TEST_F(StringsConvertTest, ToInteger) "-2147483648", "2147483648"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto results = cudf::strings::to_integers(cudf::strings_column_view(strings), cudf::data_type{cudf::type_id::INT16}); auto const expected_i16 = cudf::test::fixed_width_column_wrapper( std::initializer_list{0, 1234, NULL_VAL, 0, -9832, 93, 765, NULL_VAL, -1, -1, 0, 0}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_i16); results = cudf::strings::to_integers(cudf::strings_column_view(strings), @@ -205,7 +205,7 @@ TEST_F(StringsConvertTest, ToInteger) auto const expected_i32 = cudf::test::fixed_width_column_wrapper( std::initializer_list{ 0, 1234, NULL_VAL, 0, -9832, 93, 765, NULL_VAL, -1, 2147483647, -2147483648, -2147483648}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_i32); results = cudf::strings::to_integers(cudf::strings_column_view(strings), @@ -223,7 +223,7 @@ TEST_F(StringsConvertTest, ToInteger) 2147483647, 2147483648, 2147483648}, - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected_u32); } @@ -238,14 +238,14 @@ TEST_F(StringsConvertTest, FromInteger) cudf::test::fixed_width_column_wrapper integers( h_integers.begin(), h_integers.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = cudf::strings::from_integers(integers); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -328,9 +328,9 @@ TEST_F(StringsConvertTest, HexToInteger) std::vector h_strings{ "1234", nullptr, "98BEEF", "1a5", "CAFE", "2face", "0xAABBCCDD", "112233445566"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); { std::vector h_expected; @@ -346,7 +346,7 @@ TEST_F(StringsConvertTest, HexToInteger) cudf::test::fixed_width_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -363,7 +363,7 @@ TEST_F(StringsConvertTest, HexToInteger) cudf::test::fixed_width_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } } @@ -383,9 +383,9 @@ TEST_F(StringsConvertTest, IsHex) "0x", "x"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::test::fixed_width_column_wrapper expected( {0, 1, 0, 1, 1, 1, 1, 1, 0, 1, 0, 0}, {true, true, false, true, true, true, true, true, true, true, true, true}); diff --git a/cpp/tests/strings/ipv4_tests.cpp b/cpp/tests/strings/ipv4_tests.cpp index 78787631eeaa..e563ea37dc7e 100644 --- a/cpp/tests/strings/ipv4_tests.cpp +++ b/cpp/tests/strings/ipv4_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,7 +10,7 @@ #include #include -#include +#include #include @@ -23,8 +23,7 @@ TEST_F(StringsConvertTest, IPv4ToIntegers) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.begin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto const str) { return str != nullptr; })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::ipv4_to_integers(strings_view); @@ -33,8 +32,7 @@ TEST_F(StringsConvertTest, IPv4ToIntegers) cudf::test::fixed_width_column_wrapper expected( h_expected.cbegin(), h_expected.cend(), - thrust::make_transform_iterator(h_strings.begin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto const str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -45,15 +43,13 @@ TEST_F(StringsConvertTest, IntegersToIPv4) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.begin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto const str) { return str != nullptr; })); std::vector h_column{3232235521, 167772161, 0, 0, 700055553, 700776449}; cudf::test::fixed_width_column_wrapper column( h_column.cbegin(), h_column.cend(), - thrust::make_transform_iterator(h_strings.begin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.begin(), [](auto const str) { return str != nullptr; })); auto results = cudf::strings::integers_to_ipv4(column); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, strings); @@ -90,9 +86,9 @@ TEST_F(StringsConvertTest, IsIPv4) "9.1.2.3.4", "8.9"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::test::fixed_width_column_wrapper expected( {0, 0, 0, 1, 0, 0, 1, 0, 0, 0, 0, 0}, {true, true, false, true, true, true, true, true, true, true, true, true}); diff --git a/cpp/tests/strings/pad_tests.cpp b/cpp/tests/strings/pad_tests.cpp index b74c5b7d7c0d..c233dda0105c 100644 --- a/cpp/tests/strings/pad_tests.cpp +++ b/cpp/tests/strings/pad_tests.cpp @@ -15,7 +15,7 @@ #include #include -#include +#include #include @@ -25,9 +25,9 @@ TEST_F(StringsPadTest, Padding) { std::vector h_strings{"eee ddd", "bb cc", nullptr, "", "aa", "bbb", "ééé", "o"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::size_type width = 6; std::string phil = "+"; auto strings_view = cudf::strings_column_view(strings); @@ -40,7 +40,7 @@ TEST_F(StringsPadTest, Padding) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -51,7 +51,7 @@ TEST_F(StringsPadTest, Padding) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -62,7 +62,7 @@ TEST_F(StringsPadTest, Padding) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } } @@ -205,9 +205,9 @@ TEST_F(StringsPadTest, Wrap1) { std::vector h_strings{"12345", "thesé", nullptr, "ARE THE", "tést strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::size_type width = 3; auto strings_view = cudf::strings_column_view(strings); @@ -218,7 +218,7 @@ TEST_F(StringsPadTest, Wrap1) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -227,9 +227,9 @@ TEST_F(StringsPadTest, Wrap2) std::vector h_strings{"the quick brown fox jumped over the lazy brown dog", "hello, world"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::size_type width = 12; auto strings_view = cudf::strings_column_view(strings); @@ -241,7 +241,7 @@ TEST_F(StringsPadTest, Wrap2) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -249,9 +249,9 @@ TEST_F(StringsPadTest, WrapExpectFailure) { std::vector h_strings{"12345", "thesé", nullptr, "ARE THE", "tést strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::size_type width = 0; // this should trigger failure diff --git a/cpp/tests/strings/replace_regex_tests.cpp b/cpp/tests/strings/replace_regex_tests.cpp index 8422264d4c36..8380735e863b 100644 --- a/cpp/tests/strings/replace_regex_tests.cpp +++ b/cpp/tests/strings/replace_regex_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,7 +14,7 @@ #include #include -#include +#include #include #include @@ -424,9 +424,9 @@ TEST_F(StringsReplaceRegexTest, MediumReplaceRegex) "12345678901234567890", "abcdefghijklmnopqrstuvwxyz"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::replace_re(strings_view, *prog); @@ -435,7 +435,7 @@ TEST_F(StringsReplaceRegexTest, MediumReplaceRegex) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -453,9 +453,9 @@ TEST_F(StringsReplaceRegexTest, LargeReplaceRegex) "12345678901234567890", "abcdefghijklmnopqrstuvwxyz"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::replace_re(strings_view, *prog); @@ -464,7 +464,7 @@ TEST_F(StringsReplaceRegexTest, LargeReplaceRegex) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/replace_tests.cpp b/cpp/tests/strings/replace_tests.cpp index 8abbfa3af1d3..6985345d8cea 100644 --- a/cpp/tests/strings/replace_tests.cpp +++ b/cpp/tests/strings/replace_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -29,10 +29,9 @@ struct StringsReplaceTest : public cudf::test::BaseFixture { "", nullptr}; - return { - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })}; + return {h_strings.begin(), + h_strings.end(), + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })}; } std::unique_ptr build_large(cudf::column_view const& first, @@ -293,9 +292,9 @@ TEST_F(StringsReplaceTest, ReplaceSlice) std::vector h_strings{"Héllo", "thesé", nullptr, "ARE THE", "tést strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); { @@ -305,7 +304,7 @@ TEST_F(StringsReplaceTest, ReplaceSlice) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -315,7 +314,7 @@ TEST_F(StringsReplaceTest, ReplaceSlice) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } { @@ -325,7 +324,7 @@ TEST_F(StringsReplaceTest, ReplaceSlice) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } } @@ -362,7 +361,7 @@ TEST_F(StringsReplaceTest, ReplaceMulti) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -382,7 +381,7 @@ TEST_F(StringsReplaceTest, ReplaceMulti) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } } diff --git a/cpp/tests/strings/slice_tests.cpp b/cpp/tests/strings/slice_tests.cpp index 9ac2372966ee..8283aacd0b99 100644 --- a/cpp/tests/strings/slice_tests.cpp +++ b/cpp/tests/strings/slice_tests.cpp @@ -11,8 +11,8 @@ #include #include +#include #include -#include #include #include @@ -24,14 +24,14 @@ TEST_F(StringsSliceTest, Substring) { std::vector h_strings{"Héllo", "thesé", nullptr, "ARE THE", "tést strings", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector h_expected({"llo", "esé", nullptr, "E T", "st ", ""}); cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto strings_column = static_cast(strings); auto results = cudf::strings::slice_strings( @@ -127,9 +127,9 @@ TEST_P(Parameters, AllNulls) { std::vector h_strings{nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto start = std::optional(GetParam()); auto strings_column = cudf::strings_column_view(strings); @@ -137,9 +137,9 @@ TEST_P(Parameters, AllNulls) std::vector h_expected(h_strings); cudf::test::strings_column_wrapper expected( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_expected.begin(), [](auto str) { + return str != nullptr; + })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); thrust::host_vector starts(h_strings.size(), 1); diff --git a/cpp/tests/strings/split_tests.cpp b/cpp/tests/strings/split_tests.cpp index ce9fd7a8b10e..74e6e03ae3ea 100644 --- a/cpp/tests/strings/split_tests.cpp +++ b/cpp/tests/strings/split_tests.cpp @@ -18,7 +18,7 @@ #include #include -#include +#include #include @@ -29,21 +29,21 @@ TEST_F(StringsSplitTest, Split) std::vector h_strings{ "Héllo thesé", nullptr, "are some", "tést String", "", "no-delimiter"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); std::vector h_expected1{"Héllo", nullptr, "are", "tést", "", "no-delimiter"}; cudf::test::strings_column_wrapper expected1( h_expected1.begin(), h_expected1.end(), - thrust::make_transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); std::vector h_expected2{"thesé", nullptr, "some", "String", nullptr, nullptr}; cudf::test::strings_column_wrapper expected2( h_expected2.begin(), h_expected2.end(), - thrust::make_transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -119,9 +119,9 @@ TEST_F(StringsSplitTest, RSplit) std::vector h_strings{ "héllo", nullptr, "a_bc_déf", "a__bc", "_ab_cd", "ab_cd_", "", " a b ", " a bbb c"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); std::vector h_expected1{ @@ -129,19 +129,19 @@ TEST_F(StringsSplitTest, RSplit) cudf::test::strings_column_wrapper expected1( h_expected1.begin(), h_expected1.end(), - thrust::make_transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); std::vector h_expected2{ nullptr, nullptr, "bc", "", "ab", "cd", nullptr, nullptr, nullptr}; cudf::test::strings_column_wrapper expected2( h_expected2.begin(), h_expected2.end(), - thrust::make_transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); std::vector h_expected3{ nullptr, nullptr, "déf", "bc", "cd", "", nullptr, nullptr, nullptr}; cudf::test::strings_column_wrapper expected3( h_expected3.begin(), h_expected3.end(), - thrust::make_transform_iterator(h_expected3.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected3.begin(), [](auto str) { return str != nullptr; })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -176,26 +176,26 @@ TEST_F(StringsSplitTest, RSplitWhitespace) { std::vector h_strings{"héllo", nullptr, "a_bc_déf", "", " a\tb ", " a\r bbb c"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); std::vector h_expected1{"héllo", nullptr, "a_bc_déf", nullptr, "a", "a"}; cudf::test::strings_column_wrapper expected1( h_expected1.begin(), h_expected1.end(), - thrust::make_transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected1.begin(), [](auto str) { return str != nullptr; })); std::vector h_expected2{nullptr, nullptr, nullptr, nullptr, "b", "bbb"}; cudf::test::strings_column_wrapper expected2( h_expected2.begin(), h_expected2.end(), - thrust::make_transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected2.begin(), [](auto str) { return str != nullptr; })); std::vector h_expected3{nullptr, nullptr, nullptr, nullptr, nullptr, "c"}; cudf::test::strings_column_wrapper expected3( h_expected3.begin(), h_expected3.end(), - thrust::make_transform_iterator(h_expected3.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected3.begin(), [](auto str) { return str != nullptr; })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -431,7 +431,7 @@ TEST_F(StringsSplitTest, SplitRegex) { std::vector h_strings{" Héllo thesé", nullptr, "are some ", "tést String", ""}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper input(h_strings.begin(), h_strings.end(), validity); auto sv = cudf::strings_column_view(input); @@ -478,7 +478,7 @@ TEST_F(StringsSplitTest, SplitRecordRegex) { std::vector h_strings{" Héllo thesé", nullptr, "are some ", "tést String", ""}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper input(h_strings.begin(), h_strings.end(), validity); auto sv = cudf::strings_column_view(input); @@ -550,7 +550,7 @@ TEST_F(StringsSplitTest, SplitRegexWithMaxSplit) { std::vector h_strings{" Héllo\tthesé", nullptr, "are\nsome ", "tést\rString", ""}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper input(h_strings.begin(), h_strings.end(), validity); auto sv = cudf::strings_column_view(input); { @@ -648,7 +648,7 @@ TEST_F(StringsSplitTest, RSplitRecord) std::vector h_strings{ "héllo", nullptr, "a_bc_déf", "a__bc", "_ab_cd", "ab_cd_", "", " a b ", " a bbb c"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); using LCW = cudf::test::lists_column_wrapper; @@ -679,7 +679,7 @@ TEST_F(StringsSplitTest, RSplitRecordWithMaxSplit) " a b ___", "___ a bbb c"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); using LCW = cudf::test::lists_column_wrapper; @@ -704,7 +704,7 @@ TEST_F(StringsSplitTest, RSplitRecordWhitespace) { std::vector h_strings{"héllo", nullptr, "a_bc_déf", "", " a\tb ", " a\r bbb c"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); using LCW = cudf::test::lists_column_wrapper; @@ -721,7 +721,7 @@ TEST_F(StringsSplitTest, RSplitRecordWhitespaceWithMaxSplit) std::vector h_strings{ " héllo Asher ", nullptr, " a_bc_déf ", "", " a\tb ", " a\r bbb c"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); using LCW = cudf::test::lists_column_wrapper; @@ -738,7 +738,7 @@ TEST_F(StringsSplitTest, RSplitRegexWithMaxSplit) { std::vector h_strings{" Héllo\tthesé", nullptr, "are some\n ", "tést\rString", ""}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper input(h_strings.begin(), h_strings.end(), validity); auto sv = cudf::strings_column_view(input); @@ -926,28 +926,28 @@ TEST_F(StringsSplitTest, Partition) "_", "_", "", "", "", nullptr, "bc_déf", "_bc", "ab_cd", "cd_", "", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); auto results = cudf::strings::partition(strings_view, cudf::string_scalar("_")); EXPECT_TRUE(results->num_columns() == 3); auto exp_itr = h_expecteds.begin(); cudf::test::strings_column_wrapper expected1( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected2( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected3( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -966,28 +966,28 @@ TEST_F(StringsSplitTest, PartitionWhitespace) "", nullptr, "bc déf", " bc", "ab cd", "cd ", "", ""}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); auto results = cudf::strings::partition(strings_view); EXPECT_TRUE(results->num_columns() == 3); auto exp_itr = h_expecteds.begin(); cudf::test::strings_column_wrapper expected1( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected2( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected3( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -1006,28 +1006,28 @@ TEST_F(StringsSplitTest, RPartition) "héllo", nullptr, "déf", "bc", "cd", "", "", " a b "}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); auto results = cudf::strings::rpartition(strings_view, cudf::string_scalar("_")); EXPECT_TRUE(results->num_columns() == 3); auto exp_itr = h_expecteds.begin(); cudf::test::strings_column_wrapper expected1( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected2( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected3( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); @@ -1046,28 +1046,28 @@ TEST_F(StringsSplitTest, RPartitionWhitespace) "héllo", nullptr, "déf", "bc", "cd", "", "", "a_b"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); auto results = cudf::strings::rpartition(strings_view); EXPECT_TRUE(results->num_columns() == 3); auto exp_itr = h_expecteds.begin(); cudf::test::strings_column_wrapper expected1( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected2( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); exp_itr += h_strings.size(); cudf::test::strings_column_wrapper expected3( - exp_itr, - exp_itr + h_strings.size(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + exp_itr, exp_itr + h_strings.size(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::vector> expected_columns; expected_columns.push_back(expected1.release()); expected_columns.push_back(expected2.release()); diff --git a/cpp/tests/strings/strip_tests.cpp b/cpp/tests/strings/strip_tests.cpp index 000ea81a5bfa..b0a1cdeb8176 100644 --- a/cpp/tests/strings/strip_tests.cpp +++ b/cpp/tests/strings/strip_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,7 +11,7 @@ #include #include -#include +#include #include @@ -23,9 +23,9 @@ TEST_F(StringsStripTest, StripLeft) std::vector h_expected{"aBc ", "", nullptr, "aaaa ", "b", "ccc ddd"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::strip(strings_view, cudf::strings::side_type::LEFT); @@ -33,7 +33,7 @@ TEST_F(StringsStripTest, StripLeft) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -43,9 +43,9 @@ TEST_F(StringsStripTest, StripRight) std::vector h_expected{" aBc", "", nullptr, "", "b", "\tccc ddd"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = @@ -54,7 +54,7 @@ TEST_F(StringsStripTest, StripRight) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -64,9 +64,9 @@ TEST_F(StringsStripTest, StripBoth) std::vector h_expected{"aBc", "", nullptr, "", "b", "ccc ddd"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); auto results = @@ -75,7 +75,7 @@ TEST_F(StringsStripTest, StripBoth) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/strings/translate_tests.cpp b/cpp/tests/strings/translate_tests.cpp index a34b71c37815..cf6e7b4ea8c5 100644 --- a/cpp/tests/strings/translate_tests.cpp +++ b/cpp/tests/strings/translate_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include #include -#include +#include #include @@ -32,9 +32,9 @@ TEST_F(StringsTranslateTest, Translate) { std::vector h_strings{"eee ddd", "bb cc", nullptr, "", "aa", "débd"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); auto strings_view = cudf::strings_column_view(strings); std::vector> translate_table{ @@ -45,7 +45,7 @@ TEST_F(StringsTranslateTest, Translate) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -65,7 +65,7 @@ TEST_F(StringsTranslateTest, FilterCharacters) { std::vector h_strings{"eee ddd", "bb cc", nullptr, "", "12309", "débd"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); auto strings_view = cudf::strings_column_view(strings); diff --git a/cpp/tests/strings/urls_tests.cpp b/cpp/tests/strings/urls_tests.cpp index b70caf40ad0b..2cba047e7030 100644 --- a/cpp/tests/strings/urls_tests.cpp +++ b/cpp/tests/strings/urls_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2023, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,7 +10,7 @@ #include #include -#include +#include #include #include @@ -31,8 +31,7 @@ TEST_F(StringsConvertTest, UrlEncode) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.cbegin(), [](auto const str) { return str != nullptr; })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::url_encode(strings_view); @@ -49,8 +48,7 @@ TEST_F(StringsConvertTest, UrlEncode) cudf::test::strings_column_wrapper expected( h_expected.cbegin(), h_expected.cend(), - thrust::make_transform_iterator(h_expected.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.cbegin(), [](auto const str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -67,8 +65,7 @@ TEST_F(StringsConvertTest, UrlDecode) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.cbegin(), [](auto const str) { return str != nullptr; })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::url_decode(strings_view); @@ -84,8 +81,7 @@ TEST_F(StringsConvertTest, UrlDecode) cudf::test::strings_column_wrapper expected( h_expected.cbegin(), h_expected.cend(), - thrust::make_transform_iterator(h_expected.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.cbegin(), [](auto const str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } @@ -101,8 +97,7 @@ TEST_F(StringsConvertTest, UrlDecodeNop) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.cbegin(), [](auto const str) { return str != nullptr; })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::url_decode(strings_view); @@ -123,8 +118,7 @@ TEST_F(StringsConvertTest, UrlDecodeSliced) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.cbegin(), [](auto const str) { return str != nullptr; })); std::vector h_expected{"www.nvidia.com/rapids/?p=é%", "01/_file-1234567890.txt", @@ -137,8 +131,7 @@ TEST_F(StringsConvertTest, UrlDecodeSliced) cudf::test::strings_column_wrapper expected( h_expected.cbegin(), h_expected.cend(), - thrust::make_transform_iterator(h_expected.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.cbegin(), [](auto const str) { return str != nullptr; })); std::vector slice_indices{0, 3, 3, 6, 6, 8}; auto sliced_strings = cudf::slice(strings, slice_indices); @@ -197,8 +190,7 @@ TEST_F(StringsConvertTest, UrlDecodeLargeStrings) cudf::test::strings_column_wrapper strings( h_strings.cbegin(), h_strings.cend(), - thrust::make_transform_iterator(h_strings.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_strings.cbegin(), [](auto const str) { return str != nullptr; })); auto strings_view = cudf::strings_column_view(strings); auto results = cudf::strings::url_decode(strings_view); @@ -207,8 +199,7 @@ TEST_F(StringsConvertTest, UrlDecodeLargeStrings) cudf::test::strings_column_wrapper expected( h_expected.cbegin(), h_expected.cend(), - thrust::make_transform_iterator(h_expected.cbegin(), - [](auto const str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.cbegin(), [](auto const str) { return str != nullptr; })); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); } diff --git a/cpp/tests/text/ngrams_tests.cpp b/cpp/tests/text/ngrams_tests.cpp index 9fec3e8c29d7..f072c97f6fe6 100644 --- a/cpp/tests/text/ngrams_tests.cpp +++ b/cpp/tests/text/ngrams_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -15,7 +15,7 @@ #include -#include +#include struct TextGenerateNgramsTest : public cudf::test::BaseFixture {}; diff --git a/cpp/tests/text/ngrams_tokenize_tests.cpp b/cpp/tests/text/ngrams_tokenize_tests.cpp index 27d1288ce25a..e6ada1921a0a 100644 --- a/cpp/tests/text/ngrams_tokenize_tests.cpp +++ b/cpp/tests/text/ngrams_tokenize_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include -#include +#include #include @@ -28,9 +28,9 @@ TEST_F(TextNgramsTokenizeTest, Tokenize) "", "the mousé ate the cheese"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); diff --git a/cpp/tests/text/normalize_tests.cpp b/cpp/tests/text/normalize_tests.cpp index 050f7e83eafe..af3fb23dbf18 100644 --- a/cpp/tests/text/normalize_tests.cpp +++ b/cpp/tests/text/normalize_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include -#include +#include #include @@ -30,9 +30,9 @@ TEST_F(TextNormalizeTest, NormalizeSpaces) "no change", "the mousé ate the cheese"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); @@ -47,7 +47,7 @@ TEST_F(TextNormalizeTest, NormalizeSpaces) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto const results = nvtext::normalize_spaces(strings_view); CUDF_TEST_EXPECT_COLUMNS_EQUAL(*results, expected); diff --git a/cpp/tests/text/replace_tests.cpp b/cpp/tests/text/replace_tests.cpp index 1c5952e62570..da95e17d0de3 100644 --- a/cpp/tests/text/replace_tests.cpp +++ b/cpp/tests/text/replace_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -13,7 +13,7 @@ #include -#include +#include #include @@ -28,9 +28,9 @@ TEST_F(TextReplaceTest, ReplaceTokens) "no change", "thé is the cheese is"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::test::strings_column_wrapper targets({"is", "the"}); cudf::test::strings_column_wrapper repls({"___", ""}); std::vector h_expected{" fox jumped over dog", @@ -42,7 +42,7 @@ TEST_F(TextReplaceTest, ReplaceTokens) cudf::test::strings_column_wrapper expected( h_expected.begin(), h_expected.end(), - thrust::make_transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); + cuda::transform_iterator(h_expected.begin(), [](auto str) { return str != nullptr; })); auto results = nvtext::replace_tokens(cudf::strings_column_view(strings), cudf::strings_column_view(targets), diff --git a/cpp/tests/text/stemmer_tests.cpp b/cpp/tests/text/stemmer_tests.cpp index 184754656862..13d7f5c69aaa 100644 --- a/cpp/tests/text/stemmer_tests.cpp +++ b/cpp/tests/text/stemmer_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,7 +12,7 @@ #include -#include +#include #include @@ -35,7 +35,7 @@ TEST_F(TextStemmerTest, PorterStemmer) "private", "orrery"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); cudf::test::fixed_width_column_wrapper expected( @@ -61,7 +61,7 @@ TEST_F(TextStemmerTest, IsLetterIndex) "private", "orrery"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); cudf::strings_column_view sv(strings); @@ -120,7 +120,7 @@ TEST_F(TextStemmerTest, IsLetterIndices) "private", "orrery"}; auto validity = - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }); cudf::test::strings_column_wrapper strings(h_strings.begin(), h_strings.end(), validity); cudf::test::fixed_width_column_wrapper indices( {0, 1, 2, 3, 4, 5, 4, 3, 2, 1, -1, -2, -3, -4}); diff --git a/cpp/tests/text/tokenize_tests.cpp b/cpp/tests/text/tokenize_tests.cpp index 5dc7ec4f3162..0da7a87e2825 100644 --- a/cpp/tests/text/tokenize_tests.cpp +++ b/cpp/tests/text/tokenize_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,7 +14,7 @@ #include -#include +#include #include @@ -29,9 +29,9 @@ TEST_F(TextTokenizeTest, Tokenize) "", "the mousé ate the cheese"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); @@ -61,9 +61,9 @@ TEST_F(TextTokenizeTest, TokenizeMulti) "the over ", "the mousé ate the cheese"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); cudf::strings_column_view strings_view(strings); cudf::test::strings_column_wrapper delimiters{"the ", "over "}; diff --git a/cpp/tests/transform/one_hot_encode_tests.cpp b/cpp/tests/transform/one_hot_encode_tests.cpp index 72574afa9d0a..fb35743938da 100644 --- a/cpp/tests/transform/one_hot_encode_tests.cpp +++ b/cpp/tests/transform/one_hot_encode_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ diff --git a/cpp/tests/utilities/column_utilities.cu b/cpp/tests/utilities/column_utilities.cu index 862ac9372c04..43f17dfc1bd1 100644 --- a/cpp/tests/utilities/column_utilities.cu +++ b/cpp/tests/utilities/column_utilities.cu @@ -35,7 +35,6 @@ #include #include #include -#include #include #include #include @@ -627,11 +626,11 @@ struct column_comparator_impl { // left side size_type lhs_shift = cudf::detail::get_value(lhs_l.offsets(), lhs_l.offset(), stream); - auto lhs_offsets = thrust::make_transform_iterator( + auto lhs_offsets = cuda::transform_iterator( lhs_l.offsets().begin() + lhs_l.offset(), cuda::proclaim_return_type( [lhs_shift] __device__(size_type offset) { return offset - lhs_shift; })); - auto lhs_valids = thrust::make_transform_iterator( + auto lhs_valids = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [mask = lhs_l.null_mask(), offset = lhs_l.offset()] __device__(size_type index) { @@ -641,11 +640,11 @@ struct column_comparator_impl { // right side size_type rhs_shift = cudf::detail::get_value(rhs_l.offsets(), rhs_l.offset(), stream); - auto rhs_offsets = thrust::make_transform_iterator( + auto rhs_offsets = cuda::transform_iterator( rhs_l.offsets().begin() + rhs_l.offset(), cuda::proclaim_return_type( [rhs_shift] __device__(size_type offset) { return offset - rhs_shift; })); - auto rhs_valids = thrust::make_transform_iterator( + auto rhs_valids = cuda::transform_iterator( cuda::counting_iterator{0}, cuda::proclaim_return_type( [mask = rhs_l.null_mask(), offset = rhs_l.offset()] __device__(size_type index) { @@ -1013,9 +1012,11 @@ std::pair, std::vector> to_host(column_view auto col_span = cudf::device_span(c.begin(), c.size()); auto host_rep_types = cudf::detail::make_host_vector(col_span, stream); - auto to_fp = [&](Rep val) { return T{scaled_integer{val, scale_type{c.type().scale()}}}; }; - auto begin = thrust::make_transform_iterator(std::cbegin(host_rep_types), to_fp); - auto const host_fixed_points = thrust::host_vector(begin, begin + c.size()); + auto host_fixed_points = thrust::host_vector(c.size()); + std::transform( + host_rep_types.cbegin(), host_rep_types.cend(), host_fixed_points.begin(), [&](Rep val) { + return T{scaled_integer{val, scale_type{c.type().scale()}}}; + }); return {std::move(host_fixed_points), bitmask_to_host(c, stream, mr)}; } diff --git a/cpp/tests/utilities_tests/batched_memcpy_tests.cu b/cpp/tests/utilities_tests/batched_memcpy_tests.cu index 22c6360b1d3a..ed01aeb98402 100644 --- a/cpp/tests/utilities_tests/batched_memcpy_tests.cu +++ b/cpp/tests/utilities_tests/batched_memcpy_tests.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -17,7 +17,6 @@ #include #include -#include #include #include diff --git a/cpp/tests/utilities_tests/batched_memset_tests.cu b/cpp/tests/utilities_tests/batched_memset_tests.cu index 8a345a27e191..0e2b6d8e5543 100644 --- a/cpp/tests/utilities_tests/batched_memset_tests.cu +++ b/cpp/tests/utilities_tests/batched_memset_tests.cu @@ -17,7 +17,6 @@ #include #include -#include #include diff --git a/cpp/tests/utilities_tests/column_debug_tests.cpp b/cpp/tests/utilities_tests/column_debug_tests.cpp index 866744e4263d..c08e926a1c38 100644 --- a/cpp/tests/utilities_tests/column_debug_tests.cpp +++ b/cpp/tests/utilities_tests/column_debug_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -8,7 +8,7 @@ #include #include -#include +#include #include @@ -29,7 +29,7 @@ TYPED_TEST(ColumnDebugTestIntegral, PrintColumnNumeric) std::stringstream tmp; auto string_iter = - thrust::make_transform_iterator(std::begin(std_col), [](auto e) { return std::to_string(e); }); + cuda::transform_iterator(std::begin(std_col), [](auto e) { return std::to_string(e); }); std::copy(string_iter, string_iter + std_col.size() - 1, @@ -102,9 +102,9 @@ TEST_F(ColumnDebugStringsTest, StringsToString) std::vector h_strings{"eee", "bb", nullptr, "", "aa", "bbb", "ééé"}; cudf::test::strings_column_wrapper strings( - h_strings.begin(), - h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); + h_strings.begin(), h_strings.end(), cuda::transform_iterator(h_strings.begin(), [](auto str) { + return str != nullptr; + })); std::ostringstream tmp; tmp << h_strings[0] << delimiter << h_strings[1] << delimiter << "NULL" << delimiter diff --git a/cpp/tests/utilities_tests/column_utilities_tests.cpp b/cpp/tests/utilities_tests/column_utilities_tests.cpp index ae55a56948bc..b986069f8d4b 100644 --- a/cpp/tests/utilities_tests/column_utilities_tests.cpp +++ b/cpp/tests/utilities_tests/column_utilities_tests.cpp @@ -18,7 +18,6 @@ #include #include -#include #include @@ -233,7 +232,7 @@ TEST_F(ColumnUtilitiesStringsTest, StringsToHost) cudf::test::strings_column_wrapper strings( h_strings.begin(), h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), stream, mr); auto host_data = cudf::test::to_host(strings, stream, mr); @@ -252,7 +251,7 @@ TEST_F(ColumnUtilitiesStringsTest, StringsToHostAllNulls) cudf::test::strings_column_wrapper strings( h_strings.begin(), h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), + cuda::transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), stream, mr); auto host_data = cudf::test::to_host(strings, stream, mr); diff --git a/java/src/main/native/src/aggregation128_utils.cu b/java/src/main/native/src/aggregation128_utils.cu index 252365f5557a..2daa0622f2be 100644 --- a/java/src/main/native/src/aggregation128_utils.cu +++ b/java/src/main/native/src/aggregation128_utils.cu @@ -12,7 +12,6 @@ #include #include -#include #include #include @@ -80,14 +79,13 @@ std::unique_ptr extract_chunk32(cudf::column_view const& in_col, auto const num_rows = in_col.size(); auto out_col = cudf::make_fixed_width_column(type, num_rows, copy_bitmask(in_col), in_col.null_count()); + if (num_rows == 0) { return out_col; } + auto out_view = out_col->mutable_view(); auto const in_begin = in_col.begin(); // Build an iterator for every fourth 32-bit value, i.e.: one "chunk" of a __int128_t value - thrust::transform_iterator transform_iter{ - cuda::counting_iterator{cudf::size_type{0}}, - cuda::proclaim_return_type([] __device__(auto i) { return i * 4; })}; - cuda::permutation_iterator stride_iter{in_begin + chunk_idx, transform_iter}; + cuda::strided_iterator stride_iter{in_begin + chunk_idx, cudf::size_type{4}}; thrust::copy( rmm::exec_policy_nosync(stream), stride_iter, stride_iter + num_rows, out_view.data());