From 7239a6df59009750c2308e33b76f6fa810ee0205 Mon Sep 17 00:00:00 2001 From: Conor Hoekstra Date: Tue, 23 Nov 2021 23:04:04 -0500 Subject: [PATCH] More CTAD --- cpp/include/cudf/strings/detail/gather.cuh | 2 +- cpp/include/cudf_test/column_wrapper.hpp | 7 ++----- cpp/src/copying/concatenate.cu | 2 +- cpp/src/groupby/sort/group_rank_scan.cu | 2 +- cpp/src/groupby/sort/group_scan_util.cuh | 6 +++--- .../sort/group_single_pass_reduction_util.cuh | 12 ++++++------ cpp/src/groupby/sort/group_tdigest.cu | 10 +++++----- cpp/src/quantiles/tdigest/tdigest.cu | 7 ++----- cpp/src/rolling/grouped_rolling.cu | 6 +++--- cpp/src/rolling/rolling_collect_list.cu | 2 +- cpp/src/sort/rank.cu | 2 +- cpp/src/strings/copying/concatenate.cu | 2 +- cpp/src/strings/findall.cu | 7 ++----- cpp/src/strings/repeat_strings.cu | 2 +- cpp/src/strings/split/split.cu | 14 ++++---------- cpp/tests/iterator/iterator_tests.cuh | 11 +++-------- 16 files changed, 37 insertions(+), 57 deletions(-) diff --git a/cpp/include/cudf/strings/detail/gather.cuh b/cpp/include/cudf/strings/detail/gather.cuh index ec4a88a0e46..eb7258830ce 100644 --- a/cpp/include/cudf/strings/detail/gather.cuh +++ b/cpp/include/cudf/strings/detail/gather.cuh @@ -315,7 +315,7 @@ std::unique_ptr gather( d_out_offsets + output_count, [] __device__(auto size) { return static_cast(size); }, size_t{0}, - thrust::plus{}); + thrust::plus{}); CUDF_EXPECTS(total_bytes < static_cast(std::numeric_limits::max()), "total size of output strings is too large for a cudf column"); diff --git a/cpp/include/cudf_test/column_wrapper.hpp b/cpp/include/cudf_test/column_wrapper.hpp index e81a849b8e7..4d0c68f9ecd 100644 --- a/cpp/include/cudf_test/column_wrapper.hpp +++ b/cpp/include/cudf_test/column_wrapper.hpp @@ -1499,11 +1499,8 @@ class lists_column_wrapper : public detail::column_wrapper { // concatenate them together, skipping children that are null. std::vector children; - thrust::copy_if(std::cbegin(cols), - std::cend(cols), - valids, // stencil - std::back_inserter(children), - thrust::identity{}); + thrust::copy_if( + std::cbegin(cols), std::cend(cols), valids, std::back_inserter(children), thrust::identity{}); auto data = children.empty() ? cudf::empty_like(expected_hierarchy) : concatenate(children); diff --git a/cpp/src/copying/concatenate.cu b/cpp/src/copying/concatenate.cu index f4b6a8bf5fd..34c0cea683e 100644 --- a/cpp/src/copying/concatenate.cu +++ b/cpp/src/copying/concatenate.cu @@ -79,7 +79,7 @@ auto create_device_views(host_span views, rmm::cuda_stream_vi device_views.cend(), std::next(offsets.begin()), [](auto const& col) { return col.size(); }, - thrust::plus{}); + thrust::plus{}); auto d_offsets = make_device_uvector_async(offsets, stream); auto const output_size = offsets.back(); diff --git a/cpp/src/groupby/sort/group_rank_scan.cu b/cpp/src/groupby/sort/group_rank_scan.cu index 935ef9554a9..f36bdc0a660 100644 --- a/cpp/src/groupby/sort/group_rank_scan.cu +++ b/cpp/src/groupby/sort/group_rank_scan.cu @@ -79,7 +79,7 @@ std::unique_ptr rank_generator(column_view const& order_by, group_labels.end(), mutable_ranks.begin(), mutable_ranks.begin(), - thrust::equal_to{}, + thrust::equal_to{}, scan_op); return ranks; diff --git a/cpp/src/groupby/sort/group_scan_util.cuh b/cpp/src/groupby/sort/group_scan_util.cuh index b565e8dc6d8..15afe158af1 100644 --- a/cpp/src/groupby/sort/group_scan_util.cuh +++ b/cpp/src/groupby/sort/group_scan_util.cuh @@ -115,7 +115,7 @@ struct group_scan_functor() group_labels.end(), inp_iter, out_iter, - thrust::equal_to{}, + thrust::equal_to{}, binop); }; @@ -160,7 +160,7 @@ struct group_scan_functor{}, + thrust::equal_to{}, binop); }; @@ -214,7 +214,7 @@ struct group_scan_functor{}, + thrust::equal_to{}, binop); }; diff --git a/cpp/src/groupby/sort/group_single_pass_reduction_util.cuh b/cpp/src/groupby/sort/group_single_pass_reduction_util.cuh index c41fab144a8..95a36f40e57 100644 --- a/cpp/src/groupby/sort/group_single_pass_reduction_util.cuh +++ b/cpp/src/groupby/sort/group_single_pass_reduction_util.cuh @@ -191,7 +191,7 @@ struct group_reduction_functor{}, + thrust::equal_to{}, binop); }; @@ -215,7 +215,7 @@ struct group_reduction_functor validity(num_groups, stream); do_reduction(cudf::detail::make_validity_iterator(*d_values_ptr), validity.begin(), - thrust::logical_or{}); + thrust::logical_or{}); auto [null_mask, null_count] = cudf::detail::valid_if(validity.begin(), validity.end(), thrust::identity{}, stream, mr); @@ -264,7 +264,7 @@ struct group_reduction_functor< inp_iter, thrust::make_discard_iterator(), out_iter, - thrust::equal_to{}, + thrust::equal_to{}, binop); }; @@ -283,10 +283,10 @@ struct group_reduction_functor< auto validity = rmm::device_uvector(num_groups, stream); do_reduction(cudf::detail::make_validity_iterator(*d_values_ptr), validity.begin(), - thrust::logical_or{}); + thrust::logical_or{}); - auto [null_mask, null_count] = cudf::detail::valid_if( - validity.begin(), validity.end(), thrust::identity{}, stream, mr); + auto [null_mask, null_count] = + cudf::detail::valid_if(validity.begin(), validity.end(), thrust::identity{}, stream, mr); result->set_null_mask(std::move(null_mask), null_count); } else { auto const binop = diff --git a/cpp/src/groupby/sort/group_tdigest.cu b/cpp/src/groupby/sort/group_tdigest.cu index 146a6a8c31c..551eb128231 100644 --- a/cpp/src/groupby/sort/group_tdigest.cu +++ b/cpp/src/groupby/sort/group_tdigest.cu @@ -625,7 +625,7 @@ std::unique_ptr compute_tdigests(int delta, centroids_begin, // values thrust::make_discard_iterator(), // key output output, // output - thrust::equal_to{}, // key equality check + thrust::equal_to{}, // key equality check merge_centroids{}); // create final tdigest column @@ -850,8 +850,8 @@ std::unique_ptr group_merge_tdigest(column_view const& input, min_iter, thrust::make_discard_iterator(), merged_min_col->mutable_view().begin(), - thrust::equal_to{}, // key equality check - thrust::minimum{}); + thrust::equal_to{}, // key equality check + thrust::minimum{}); auto merged_max_col = cudf::make_numeric_column( data_type{type_id::FLOAT64}, num_groups, mask_state::UNALLOCATED, stream, mr); @@ -864,8 +864,8 @@ std::unique_ptr group_merge_tdigest(column_view const& input, max_iter, thrust::make_discard_iterator(), merged_max_col->mutable_view().begin(), - thrust::equal_to{}, // key equality check - thrust::maximum{}); + thrust::equal_to{}, // key equality check + thrust::maximum{}); // for any empty groups, set the min and max to be 0. not technically necessary but it makes // testing simpler. diff --git a/cpp/src/quantiles/tdigest/tdigest.cu b/cpp/src/quantiles/tdigest/tdigest.cu index 57c221b15ed..18e7d02d086 100644 --- a/cpp/src/quantiles/tdigest/tdigest.cu +++ b/cpp/src/quantiles/tdigest/tdigest.cu @@ -348,11 +348,8 @@ std::unique_ptr percentile_approx(tdigest_column_view const& input, if (null_count == 0) { return std::pair{rmm::device_buffer{}, null_count}; } - return cudf::detail::valid_if(tdigest_is_empty, - tdigest_is_empty + tdv.size(), - thrust::logical_not{}, - stream, - mr); + return cudf::detail::valid_if( + tdigest_is_empty, tdigest_is_empty + tdv.size(), thrust::logical_not{}, stream, mr); }(); return cudf::make_lists_column( diff --git a/cpp/src/rolling/grouped_rolling.cu b/cpp/src/rolling/grouped_rolling.cu index 509f67bb5c6..5a7f15148d8 100644 --- a/cpp/src/rolling/grouped_rolling.cu +++ b/cpp/src/rolling/grouped_rolling.cu @@ -142,8 +142,8 @@ std::unique_ptr grouped_rolling_window(table_view const& group_keys, preceding_window] __device__(size_type idx) { auto group_label = d_group_labels[idx]; auto group_start = d_group_offsets[group_label]; - return thrust::minimum{}(preceding_window, - idx - group_start + 1); // Preceding includes current row. + return thrust::minimum{}(preceding_window, + idx - group_start + 1); // Preceding includes current row. }; auto following_calculator = [d_group_offsets = group_offsets.data(), @@ -152,7 +152,7 @@ std::unique_ptr grouped_rolling_window(table_view const& group_keys, auto group_label = d_group_labels[idx]; auto group_end = d_group_offsets[group_label + 1]; // Cannot fall off the end, since offsets // is capped with `input.size()`. - return thrust::minimum{}(following_window, (group_end - 1) - idx); + return thrust::minimum{}(following_window, (group_end - 1) - idx); }; if (aggr.kind == aggregation::CUDA || aggr.kind == aggregation::PTX) { diff --git a/cpp/src/rolling/rolling_collect_list.cu b/cpp/src/rolling/rolling_collect_list.cu index ecef90dc8e1..30c39bde7d2 100644 --- a/cpp/src/rolling/rolling_collect_list.cu +++ b/cpp/src/rolling/rolling_collect_list.cu @@ -75,7 +75,7 @@ std::unique_ptr get_list_child_to_list_row_mapping(cudf::column_view con per_row_mapping_begin, per_row_mapping_begin + num_child_rows, per_row_mapping_begin, - thrust::maximum{}); + thrust::maximum{}); return per_row_mapping; } diff --git a/cpp/src/sort/rank.cu b/cpp/src/sort/rank.cu index cbc2d0a0730..e9589e6c4b3 100644 --- a/cpp/src/sort/rank.cu +++ b/cpp/src/sort/rank.cu @@ -117,7 +117,7 @@ void tie_break_ranks_transform(cudf::device_span dense_rank_sor tie_iter, thrust::make_discard_iterator(), tie_sorted.begin(), - thrust::equal_to{}, + thrust::equal_to{}, tie_breaker); auto sorted_tied_rank = thrust::make_transform_iterator( dense_rank_sorted.begin(), diff --git a/cpp/src/strings/copying/concatenate.cu b/cpp/src/strings/copying/concatenate.cu index db8b37a9592..3822fa8bf5a 100644 --- a/cpp/src/strings/copying/concatenate.cu +++ b/cpp/src/strings/copying/concatenate.cu @@ -96,7 +96,7 @@ auto create_strings_device_views(host_span views, rmm::cuda_s device_views_ptr + views.size(), std::next(d_partition_offsets.begin()), chars_size_transform{}, - thrust::plus{}); + thrust::plus{}); auto const output_chars_size = d_partition_offsets.back_element(stream); stream.synchronize(); // ensure copy of output_chars_size is complete before returning diff --git a/cpp/src/strings/findall.cu b/cpp/src/strings/findall.cu index 3ab5b55020c..8d96f0de415 100644 --- a/cpp/src/strings/findall.cu +++ b/cpp/src/strings/findall.cu @@ -153,11 +153,8 @@ std::unique_ptr findall_re( std::vector> results; - size_type const columns = thrust::reduce(rmm::exec_policy(stream), - find_counts.begin(), - find_counts.end(), - 0, - thrust::maximum{}); + size_type const columns = thrust::reduce( + rmm::exec_policy(stream), find_counts.begin(), find_counts.end(), 0, thrust::maximum{}); // boundary case: if no columns, return all nulls column (issue #119) if (columns == 0) results.emplace_back(std::make_unique( diff --git a/cpp/src/strings/repeat_strings.cu b/cpp/src/strings/repeat_strings.cu index 458f3ed885c..7820e0064a6 100644 --- a/cpp/src/strings/repeat_strings.cu +++ b/cpp/src/strings/repeat_strings.cu @@ -369,7 +369,7 @@ std::pair, int64_t> repeat_strings_output_sizes( thrust::make_counting_iterator(strings_count), fn, int64_t{0}, - thrust::plus{}); + thrust::plus{}); return std::make_pair(std::move(output_sizes), total_bytes); } diff --git a/cpp/src/strings/split/split.cu b/cpp/src/strings/split/split.cu index 5113b418501..c6e52a79059 100644 --- a/cpp/src/strings/split/split.cu +++ b/cpp/src/strings/split/split.cu @@ -490,11 +490,8 @@ std::unique_ptr
split_fn(strings_column_view const& strings_column, }); // the columns_count is the maximum number of tokens for any string - auto const columns_count = thrust::reduce(rmm::exec_policy(stream), - token_counts.begin(), - token_counts.end(), - 0, - thrust::maximum{}); + auto const columns_count = thrust::reduce( + rmm::exec_policy(stream), token_counts.begin(), token_counts.end(), 0, thrust::maximum{}); // boundary case: if no columns, return one null column (custrings issue #119) if (columns_count == 0) { results.push_back(std::make_unique( @@ -748,11 +745,8 @@ std::unique_ptr
whitespace_split_fn(size_type strings_count, [tokenizer] __device__(size_type idx) { return tokenizer.count_tokens(idx); }); // column count is the maximum number of tokens for any string - size_type const columns_count = thrust::reduce(rmm::exec_policy(stream), - token_counts.begin(), - token_counts.end(), - 0, - thrust::maximum{}); + size_type const columns_count = thrust::reduce( + rmm::exec_policy(stream), token_counts.begin(), token_counts.end(), 0, thrust::maximum{}); std::vector> results; // boundary case: if no columns, return one null column (issue #119) diff --git a/cpp/tests/iterator/iterator_tests.cuh b/cpp/tests/iterator/iterator_tests.cuh index 4ec347c4bc1..0cf627407e9 100644 --- a/cpp/tests/iterator/iterator_tests.cuh +++ b/cpp/tests/iterator/iterator_tests.cuh @@ -50,13 +50,8 @@ struct IteratorTest : public cudf::test::BaseFixture { // Get temporary storage size size_t temp_storage_bytes = 0; - cub::DeviceReduce::Reduce(nullptr, - temp_storage_bytes, - d_in, - dev_result.begin(), - num_items, - thrust::minimum{}, - init); + cub::DeviceReduce::Reduce( + nullptr, temp_storage_bytes, d_in, dev_result.begin(), num_items, thrust::minimum{}, init); // Allocate temporary storage rmm::device_buffer d_temp_storage(temp_storage_bytes, rmm::cuda_stream_default); @@ -67,7 +62,7 @@ struct IteratorTest : public cudf::test::BaseFixture { d_in, dev_result.begin(), num_items, - thrust::minimum{}, + thrust::minimum{}, init); evaluate(expected, dev_result, "cub test");