From 4de7e7ac204572abf4bbdcc0553a49adafc5b2c4 Mon Sep 17 00:00:00 2001 From: Shruti Shivakumar Date: Sat, 30 Nov 2024 23:41:00 -0500 Subject: [PATCH 1/4] Expose stream-ordering in replace API (#17436) Adds stream parameter to ``` cudf::normalize_nans_and_zeros ``` Reference: https://github.com/rapidsai/cudf/issues/13744 Authors: - Shruti Shivakumar (https://github.com/shrshi) Approvers: - MithunR (https://github.com/mythrocks) - Nghia Truong (https://github.com/ttnghia) URL: https://github.com/rapidsai/cudf/pull/17436 --- cpp/src/replace/nans.cu | 2 +- cpp/tests/streams/replace_test.cpp | 4 ++-- 2 files changed, 3 insertions(+), 3 deletions(-) diff --git a/cpp/src/replace/nans.cu b/cpp/src/replace/nans.cu index 394c2a2de80..2fc6da23991 100644 --- a/cpp/src/replace/nans.cu +++ b/cpp/src/replace/nans.cu @@ -244,7 +244,7 @@ std::unique_ptr normalize_nans_and_zeros(column_view const& input, void normalize_nans_and_zeros(mutable_column_view& in_out, rmm::cuda_stream_view stream) { CUDF_FUNC_RANGE(); - detail::normalize_nans_and_zeros(in_out, cudf::get_default_stream()); + detail::normalize_nans_and_zeros(in_out, stream); } } // namespace cudf diff --git a/cpp/tests/streams/replace_test.cpp b/cpp/tests/streams/replace_test.cpp index 752ddc87dfc..89f76237de6 100644 --- a/cpp/tests/streams/replace_test.cpp +++ b/cpp/tests/streams/replace_test.cpp @@ -107,6 +107,6 @@ TEST_F(ReplaceTest, NormalizeNansAndZerosMutable) auto nan = std::numeric_limits::quiet_NaN(); auto input_column = cudf::test::make_type_param_vector({-0.0, 0.0, -nan, nan, nan}); cudf::test::fixed_width_column_wrapper input(input_column.begin(), input_column.end()); - cudf::normalize_nans_and_zeros(static_cast(input), - cudf::test::get_default_stream()); + cudf::mutable_column_view mutable_view = cudf::column(input, cudf::test::get_default_stream()); + cudf::normalize_nans_and_zeros(mutable_view, cudf::test::get_default_stream()); } From 4b2dc3334c1c8597cd4e36ecf345f29b225dfd35 Mon Sep 17 00:00:00 2001 From: James Lamb Date: Mon, 2 Dec 2024 08:18:42 -0600 Subject: [PATCH 2/4] skip most CI on devcontainer-only changes (#17465) The contents of the `.devcontainer/` directory don't affect any of the artifacts produced from this repo. This proposes skipping most CI on PRs that only make changes in that directory. Authors: - James Lamb (https://github.com/jameslamb) Approvers: - Bradley Dice (https://github.com/bdice) URL: https://github.com/rapidsai/cudf/pull/17465 --- .github/workflows/pr.yaml | 5 +++++ 1 file changed, 5 insertions(+) diff --git a/.github/workflows/pr.yaml b/.github/workflows/pr.yaml index a8afede4821..7c0bd6d52e2 100644 --- a/.github/workflows/pr.yaml +++ b/.github/workflows/pr.yaml @@ -61,6 +61,7 @@ jobs: files_yaml: | test_cpp: - '**' + - '!.devcontainer/**' - '!CONTRIBUTING.md' - '!README.md' - '!ci/cudf_pandas_scripts/**' @@ -71,6 +72,7 @@ jobs: - '!python/**' test_cudf_pandas: - '**' + - '!.devcontainer/**' - '!CONTRIBUTING.md' - '!README.md' - '!docs/**' @@ -79,6 +81,7 @@ jobs: - '!notebooks/**' test_java: - '**' + - '!.devcontainer/**' - '!CONTRIBUTING.md' - '!README.md' - '!ci/cudf_pandas_scripts/**' @@ -88,12 +91,14 @@ jobs: - '!python/**' test_notebooks: - '**' + - '!.devcontainer/**' - '!CONTRIBUTING.md' - '!README.md' - '!ci/cudf_pandas_scripts/**' - '!java/**' test_python: - '**' + - '!.devcontainer/**' - '!CONTRIBUTING.md' - '!README.md' - '!ci/cudf_pandas_scripts/**' From 3e418dd05d4f84472bca4d80902e1b7476f0e0d4 Mon Sep 17 00:00:00 2001 From: David Wendt <45795991+davidwendt@users.noreply.github.com> Date: Mon, 2 Dec 2024 13:13:16 -0500 Subject: [PATCH 3/4] Move make_strings_column benchmark to nvbench (#17340) Moves the `cpp/benchmarks/string/factory.cu` implementation from google-bench to nvbench. Also renames to `.cpp` by recoding without device code. Authors: - David Wendt (https://github.com/davidwendt) Approvers: - Bradley Dice (https://github.com/bdice) - Vyas Ramasubramani (https://github.com/vyasr) URL: https://github.com/rapidsai/cudf/pull/17340 --- cpp/benchmarks/CMakeLists.txt | 3 +- cpp/benchmarks/string/factory.cpp | 60 ++++++++++++++++++++ cpp/benchmarks/string/factory.cu | 92 ------------------------------- 3 files changed, 61 insertions(+), 94 deletions(-) create mode 100644 cpp/benchmarks/string/factory.cpp delete mode 100644 cpp/benchmarks/string/factory.cu diff --git a/cpp/benchmarks/CMakeLists.txt b/cpp/benchmarks/CMakeLists.txt index d3de9b39977..8e5ea900efa 100644 --- a/cpp/benchmarks/CMakeLists.txt +++ b/cpp/benchmarks/CMakeLists.txt @@ -360,8 +360,6 @@ ConfigureNVBench( # ################################################################################################## # * strings benchmark ------------------------------------------------------------------- -ConfigureBench(STRINGS_BENCH string/factory.cu) - ConfigureNVBench( STRINGS_NVBENCH string/case.cpp @@ -377,6 +375,7 @@ ConfigureNVBench( string/copy_range.cpp string/count.cpp string/extract.cpp + string/factory.cpp string/filter.cpp string/find.cpp string/find_multiple.cpp diff --git a/cpp/benchmarks/string/factory.cpp b/cpp/benchmarks/string/factory.cpp new file mode 100644 index 00000000000..03870b0ae23 --- /dev/null +++ b/cpp/benchmarks/string/factory.cpp @@ -0,0 +1,60 @@ +/* + * Copyright (c) 2021-2024, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include + +#include +#include +#include +#include +#include + +#include + +#include + +#include + +static void bench_factory(nvbench::state& state) +{ + auto const num_rows = static_cast(state.get_int64("num_rows")); + auto const min_width = static_cast(state.get_int64("min_width")); + auto const max_width = static_cast(state.get_int64("max_width")); + + data_profile const profile = data_profile_builder().distribution( + cudf::type_id::STRING, distribution_id::NORMAL, min_width, max_width); + auto const column = create_random_column(cudf::type_id::STRING, row_count{num_rows}, profile); + auto const sv = cudf::strings_column_view(column->view()); + + auto stream = cudf::get_default_stream(); + auto mr = cudf::get_current_device_resource_ref(); + auto d_strings = cudf::strings::detail::create_string_vector_from_column(sv, stream, mr); + + state.set_cuda_stream(nvbench::make_cuda_stream_view(stream.value())); + auto chars_size = sv.chars_size(stream); + state.add_global_memory_reads(chars_size); + state.add_global_memory_writes(chars_size); + + state.exec(nvbench::exec_tag::sync, [&](nvbench::launch& launch) { + cudf::make_strings_column(d_strings, cudf::string_view{nullptr, 0}); + }); +} + +NVBENCH_BENCH(bench_factory) + .set_name("factory") + .add_int64_axis("min_width", {0}) + .add_int64_axis("max_width", {32, 64, 128, 256}) + .add_int64_axis("num_rows", {32768, 262144, 2097152}); diff --git a/cpp/benchmarks/string/factory.cu b/cpp/benchmarks/string/factory.cu deleted file mode 100644 index c4e74c4d97e..00000000000 --- a/cpp/benchmarks/string/factory.cu +++ /dev/null @@ -1,92 +0,0 @@ -/* - * Copyright (c) 2021-2024, NVIDIA CORPORATION. - * - * Licensed under the Apache License, Version 2.0 (the "License"); - * you may not use this file except in compliance with the License. - * You may obtain a copy of the License at - * - * http://www.apache.org/licenses/LICENSE-2.0 - * - * Unless required by applicable law or agreed to in writing, software - * distributed under the License is distributed on an "AS IS" BASIS, - * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. - * See the License for the specific language governing permissions and - * limitations under the License. - */ - -#include "string_bench_args.hpp" - -#include -#include -#include - -#include - -#include -#include -#include - -#include - -#include -#include -#include - -#include - -namespace { -using string_pair = thrust::pair; -struct string_view_to_pair { - __device__ string_pair operator()(thrust::pair const& p) - { - return (p.second) ? string_pair{p.first.data(), p.first.size_bytes()} : string_pair{nullptr, 0}; - } -}; -} // namespace - -class StringsFactory : public cudf::benchmark {}; - -static void BM_factory(benchmark::State& state) -{ - cudf::size_type const n_rows{static_cast(state.range(0))}; - cudf::size_type const max_str_length{static_cast(state.range(1))}; - data_profile const profile = data_profile_builder().distribution( - cudf::type_id::STRING, distribution_id::NORMAL, 0, max_str_length); - auto const column = create_random_column(cudf::type_id::STRING, row_count{n_rows}, profile); - auto d_column = cudf::column_device_view::create(column->view()); - rmm::device_uvector pairs(d_column->size(), cudf::get_default_stream()); - thrust::transform(thrust::device, - d_column->pair_begin(), - d_column->pair_end(), - pairs.data(), - string_view_to_pair{}); - - for (auto _ : state) { - cuda_event_timer raii(state, true, cudf::get_default_stream()); - cudf::make_strings_column(pairs, cudf::get_default_stream()); - } - - cudf::strings_column_view input(column->view()); - state.SetBytesProcessed(state.iterations() * input.chars_size(cudf::get_default_stream())); -} - -static void generate_bench_args(benchmark::internal::Benchmark* b) -{ - int const min_rows = 1 << 12; - int const max_rows = 1 << 24; - int const row_mult = 8; - int const min_rowlen = 1 << 5; - int const max_rowlen = 1 << 13; - int const len_mult = 4; - generate_string_bench_args(b, min_rows, max_rows, row_mult, min_rowlen, max_rowlen, len_mult); -} - -#define STRINGS_BENCHMARK_DEFINE(name) \ - BENCHMARK_DEFINE_F(StringsFactory, name) \ - (::benchmark::State & st) { BM_factory(st); } \ - BENCHMARK_REGISTER_F(StringsFactory, name) \ - ->Apply(generate_bench_args) \ - ->UseManualTime() \ - ->Unit(benchmark::kMillisecond); - -STRINGS_BENCHMARK_DEFINE(factory) From 5190b4460ba86151521de9f4415c5eb55781371e Mon Sep 17 00:00:00 2001 From: Lawrence Mitchell Date: Mon, 2 Dec 2024 18:42:47 +0000 Subject: [PATCH 4/4] Temporarily skip tests due to dask/distributed#8953 (#17472) Temporarily skip tests failing due to an upstream dask change. Authors: - Lawrence Mitchell (https://github.com/wence-) Approvers: - GALI PREM SAGAR (https://github.com/galipremsagar) - Matthew Murray (https://github.com/Matt711) URL: https://github.com/rapidsai/cudf/pull/17472 --- .../custreamz/tests/test_dataframes.py | 56 ++++++++++++++++--- 1 file changed, 49 insertions(+), 7 deletions(-) diff --git a/python/custreamz/custreamz/tests/test_dataframes.py b/python/custreamz/custreamz/tests/test_dataframes.py index 8c0130d2818..6905044039c 100644 --- a/python/custreamz/custreamz/tests/test_dataframes.py +++ b/python/custreamz/custreamz/tests/test_dataframes.py @@ -216,7 +216,13 @@ def test_set_index(): assert_eq(b[0], df.set_index(df.y + 1)) -def test_binary_stream_operators(stream): +def test_binary_stream_operators(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) df = cudf.DataFrame({"x": [1, 2, 3], "y": [4, 5, 6]}) expected = df.x + df.y @@ -242,7 +248,13 @@ def test_index(stream): assert_eq(L[1], df.index + 5) -def test_pair_arithmetic(stream): +def test_pair_arithmetic(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) df = cudf.DataFrame({"x": list(range(10)), "y": [1] * 10}) a = DataFrame(example=df.iloc[:0], stream=stream) @@ -255,7 +267,13 @@ def test_pair_arithmetic(stream): assert_eq(cudf.concat(L), (df.x + df.y) * 2) -def test_getitem(stream): +def test_getitem(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) df = cudf.DataFrame({"x": list(range(10)), "y": [1] * 10}) a = DataFrame(example=df.iloc[:0], stream=stream) @@ -332,7 +350,13 @@ def test_repr_html(stream): assert "1" in html -def test_setitem(stream): +def test_setitem(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) df = cudf.DataFrame({"x": list(range(10)), "y": [1] * 10}) sdf = DataFrame(example=df.iloc[:0], stream=stream) @@ -356,7 +380,13 @@ def test_setitem(stream): assert_eq(L[-1], df.mean()) -def test_setitem_overwrites(stream): +def test_setitem_overwrites(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) df = cudf.DataFrame({"x": list(range(10))}) sdf = DataFrame(example=df.iloc[:0], stream=stream) stream = sdf.stream @@ -413,8 +443,14 @@ def test_setitem_overwrites(stream): ], ) def test_rolling_count_aggregations( - op, window, m, pre_get, post_get, kwargs, stream + request, op, window, m, pre_get, post_get, kwargs, stream ): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream) and len(kwargs) == 0, + reason="https://github.com/dask/distributed/issues/8953", + ) + ) index = pd.DatetimeIndex( pd.date_range("2000-01-01", "2000-01-03", freq="1h") ) @@ -808,7 +844,13 @@ def test_reductions_with_start_state(stream): assert output2[0] == 360 -def test_rolling_aggs_with_start_state(stream): +def test_rolling_aggs_with_start_state(request, stream): + request.applymarker( + pytest.mark.xfail( + isinstance(stream, DaskStream), + reason="https://github.com/dask/distributed/issues/8953", + ) + ) example = cudf.DataFrame({"name": [], "amount": []}, dtype="float64") sdf = DataFrame(stream, example=example) output0 = (