From 7cd3834987c7bd635c4758b939544518cb3d1236 Mon Sep 17 00:00:00 2001 From: Suraj Aralihalli Date: Mon, 29 Jan 2024 10:34:35 -0800 Subject: [PATCH 1/3] revert sum agg (#14907) This pull request reverses the modifications made to the sum/product aggregation target type, ensuring it always produces int64. The changes implemented by PR [14679](https://github.com/rapidsai/cudf/pull/14679) which led to degraded performance when the aggregation column had an unsigned type, are reverted. Additional details can be found in the issue [14886](https://github.com/rapidsai/cudf/issues/14886). Authors: - Suraj Aralihalli (https://github.com/SurajAralihalli) Approvers: - David Wendt (https://github.com/davidwendt) - Nghia Truong (https://github.com/ttnghia) - Karthikeyan (https://github.com/karthikeyann) --- cpp/include/cudf/detail/aggregation/aggregation.hpp | 4 ++-- cpp/tests/groupby/sum_tests.cpp | 11 ++++------- 2 files changed, 6 insertions(+), 9 deletions(-) diff --git a/cpp/include/cudf/detail/aggregation/aggregation.hpp b/cpp/include/cudf/detail/aggregation/aggregation.hpp index c35d56b4c13..a8f164646a5 100644 --- a/cpp/include/cudf/detail/aggregation/aggregation.hpp +++ b/cpp/include/cudf/detail/aggregation/aggregation.hpp @@ -1234,12 +1234,12 @@ constexpr bool is_sum_product_agg(aggregation::Kind k) (k == aggregation::SUM_OF_SQUARES); } -// Summing/Multiplying integers of any type, always use uint64_t for unsigned and int64_t for signed +// Summing/Multiplying integers of any type, always use int64_t accumulator template struct target_type_impl && is_sum_product_agg(k)>> { - using type = std::conditional_t, uint64_t, int64_t>; + using type = int64_t; }; // Summing fixed_point numbers diff --git a/cpp/tests/groupby/sum_tests.cpp b/cpp/tests/groupby/sum_tests.cpp index abf25eb0aa9..03cc3fab568 100644 --- a/cpp/tests/groupby/sum_tests.cpp +++ b/cpp/tests/groupby/sum_tests.cpp @@ -28,10 +28,10 @@ using namespace cudf::test::iterators; template struct groupby_sum_test : public cudf::test::BaseFixture {}; -using K = int32_t; -using supported_types = cudf::test::Concat< - cudf::test::Types, - cudf::test::DurationTypes>; +using K = int32_t; +using supported_types = + cudf::test::Concat, + cudf::test::DurationTypes>; TYPED_TEST_SUITE(groupby_sum_test, supported_types); @@ -40,9 +40,6 @@ TYPED_TEST(groupby_sum_test, basic) using V = TypeParam; using R = cudf::detail::target_type_t; - static_assert(std::is_signed_v == std::is_signed_v, - "Both Result type and Source type must have same signedness"); - cudf::test::fixed_width_column_wrapper keys{1, 2, 3, 1, 2, 2, 1, 3, 3, 2}; cudf::test::fixed_width_column_wrapper vals{0, 1, 2, 3, 4, 5, 6, 7, 8, 9}; From 5cc021af0ef934ddf3f5f66cee2d8dd2490ba623 Mon Sep 17 00:00:00 2001 From: David Wendt <45795991+davidwendt@users.noreply.github.com> Date: Mon, 29 Jan 2024 13:51:26 -0500 Subject: [PATCH 2/3] Use offsetalator in cudf::strings::copy_slice (#14844) Replace hardcoded offset types with offsetalator in `cudf::strings::detail::copy_slice`. Authors: - David Wendt (https://github.com/davidwendt) Approvers: - Mark Harris (https://github.com/harrism) - Bradley Dice (https://github.com/bdice) URL: https://github.com/rapidsai/cudf/pull/14844 --- cpp/src/strings/copying/copying.cu | 35 ++++++++++++++++-------------- 1 file changed, 19 insertions(+), 16 deletions(-) diff --git a/cpp/src/strings/copying/copying.cu b/cpp/src/strings/copying/copying.cu index 4f37d3864ac..013028d6df3 100644 --- a/cpp/src/strings/copying/copying.cu +++ b/cpp/src/strings/copying/copying.cu @@ -16,9 +16,10 @@ #include #include -#include #include +#include #include +#include #include #include @@ -33,47 +34,49 @@ namespace cudf { namespace strings { namespace detail { -std::unique_ptr copy_slice(strings_column_view const& strings, +std::unique_ptr copy_slice(strings_column_view const& input, size_type start, size_type end, rmm::cuda_stream_view stream, rmm::mr::device_memory_resource* mr) { - if (strings.is_empty()) return make_empty_column(type_id::STRING); - if (end < 0 || end > strings.size()) end = strings.size(); + if (input.is_empty()) { return make_empty_column(type_id::STRING); } CUDF_EXPECTS(((start >= 0) && (start < end)), "Invalid start parameter value."); auto const strings_count = end - start; - auto const offsets_offset = start + strings.offset(); + auto const offsets_offset = start + input.offset(); // slice the offsets child column auto offsets_column = std::make_unique( cudf::detail::slice( - strings.offsets(), {offsets_offset, offsets_offset + strings_count + 1}, stream) + input.offsets(), {offsets_offset, offsets_offset + strings_count + 1}, stream) .front(), stream, mr); auto const chars_offset = - offsets_offset == 0 ? 0 : cudf::detail::get_value(offsets_column->view(), 0, stream); + offsets_offset == 0 ? 0L : get_offset_value(offsets_column->view(), 0, stream); if (chars_offset > 0) { // adjust the individual offset values only if needed - auto d_offsets = offsets_column->mutable_view(); + auto d_offsets = + cudf::detail::offsetalator_factory::make_output_iterator(offsets_column->mutable_view()); + auto input_offsets = + cudf::detail::offsetalator_factory::make_input_iterator(input.offsets(), offsets_offset); thrust::transform(rmm::exec_policy(stream), - d_offsets.begin(), - d_offsets.end(), - d_offsets.begin(), - cuda::proclaim_return_type( + input_offsets, + input_offsets + offsets_column->size(), + d_offsets, + cuda::proclaim_return_type( [chars_offset] __device__(auto offset) { return offset - chars_offset; })); } // slice the chars child column - auto const data_size = static_cast( - cudf::detail::get_value(offsets_column->view(), strings_count, stream)); + auto const data_size = + static_cast(get_offset_value(offsets_column->view(), strings_count, stream)); auto chars_buffer = - rmm::device_buffer{strings.chars_begin(stream) + chars_offset, data_size, stream, mr}; + rmm::device_buffer{input.chars_begin(stream) + chars_offset, data_size, stream, mr}; // slice the null mask auto null_mask = cudf::detail::copy_bitmask( - strings.null_mask(), offsets_offset, offsets_offset + strings_count, stream, mr); + input.null_mask(), offsets_offset, offsets_offset + strings_count, stream, mr); auto null_count = cudf::detail::null_count( static_cast(null_mask.data()), 0, strings_count, stream); From fc2b9771f17644243817a339e218360aa97a1a79 Mon Sep 17 00:00:00 2001 From: GALI PREM SAGAR Date: Mon, 29 Jan 2024 13:29:46 -0600 Subject: [PATCH 3/3] Pin pytest to <8 (#14920) --- conda/environments/all_cuda-118_arch-x86_64.yaml | 2 +- conda/environments/all_cuda-120_arch-x86_64.yaml | 2 +- dependencies.yaml | 2 +- python/cudf/pyproject.toml | 2 +- python/cudf_kafka/pyproject.toml | 2 +- python/custreamz/pyproject.toml | 2 +- python/dask_cudf/pyproject.toml | 2 +- 7 files changed, 7 insertions(+), 7 deletions(-) diff --git a/conda/environments/all_cuda-118_arch-x86_64.yaml b/conda/environments/all_cuda-118_arch-x86_64.yaml index 47b377013ce..956c685f7de 100644 --- a/conda/environments/all_cuda-118_arch-x86_64.yaml +++ b/conda/environments/all_cuda-118_arch-x86_64.yaml @@ -73,11 +73,11 @@ dependencies: - ptxcompiler - pyarrow==14.0.1.* - pydata-sphinx-theme!=0.14.2 -- pytest - pytest-benchmark - pytest-cases>=3.8.2 - pytest-cov - pytest-xdist +- pytest<8 - python-confluent-kafka>=1.9.0,<1.10.0a0 - python-snappy>=0.6.0 - python>=3.9,<3.11 diff --git a/conda/environments/all_cuda-120_arch-x86_64.yaml b/conda/environments/all_cuda-120_arch-x86_64.yaml index 4cf1d5427f4..cd2c70577f9 100644 --- a/conda/environments/all_cuda-120_arch-x86_64.yaml +++ b/conda/environments/all_cuda-120_arch-x86_64.yaml @@ -71,11 +71,11 @@ dependencies: - pyarrow==14.0.1.* - pydata-sphinx-theme!=0.14.2 - pynvjitlink -- pytest - pytest-benchmark - pytest-cases>=3.8.2 - pytest-cov - pytest-xdist +- pytest<8 - python-confluent-kafka>=1.9.0,<1.10.0a0 - python-snappy>=0.6.0 - python>=3.9,<3.11 diff --git a/dependencies.yaml b/dependencies.yaml index 90a04b2f876..9a1d11af02d 100644 --- a/dependencies.yaml +++ b/dependencies.yaml @@ -612,7 +612,7 @@ dependencies: common: - output_types: [conda, requirements, pyproject] packages: - - pytest + - pytest<8 - pytest-cov - pytest-xdist test_python_cudf: diff --git a/python/cudf/pyproject.toml b/python/cudf/pyproject.toml index c7b66abea27..ce30230398f 100644 --- a/python/cudf/pyproject.toml +++ b/python/cudf/pyproject.toml @@ -57,11 +57,11 @@ test = [ "fastavro>=0.22.9", "hypothesis", "msgpack", - "pytest", "pytest-benchmark", "pytest-cases>=3.8.2", "pytest-cov", "pytest-xdist", + "pytest<8", "python-snappy>=0.6.0", "scipy", "tokenizers==0.13.1", diff --git a/python/cudf_kafka/pyproject.toml b/python/cudf_kafka/pyproject.toml index 062a0224c1f..d6574c32873 100644 --- a/python/cudf_kafka/pyproject.toml +++ b/python/cudf_kafka/pyproject.toml @@ -27,9 +27,9 @@ dependencies = [ [project.optional-dependencies] test = [ - "pytest", "pytest-cov", "pytest-xdist", + "pytest<8", ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`. [project.urls] diff --git a/python/custreamz/pyproject.toml b/python/custreamz/pyproject.toml index 3e6c74ab570..03ec079a890 100644 --- a/python/custreamz/pyproject.toml +++ b/python/custreamz/pyproject.toml @@ -36,9 +36,9 @@ classifiers = [ [project.optional-dependencies] test = [ - "pytest", "pytest-cov", "pytest-xdist", + "pytest<8", ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`. [project.urls] diff --git a/python/dask_cudf/pyproject.toml b/python/dask_cudf/pyproject.toml index 33065da6e8d..c3185bcb793 100644 --- a/python/dask_cudf/pyproject.toml +++ b/python/dask_cudf/pyproject.toml @@ -43,9 +43,9 @@ cudf = "dask_cudf.backends:CudfBackendEntrypoint" test = [ "dask-cuda==24.2.*", "numba>=0.57", - "pytest", "pytest-cov", "pytest-xdist", + "pytest<8", ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`. [project.urls]