diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt index 06c0462ebed9..6a12521efb07 100644 --- a/cpp/tests/CMakeLists.txt +++ b/cpp/tests/CMakeLists.txt @@ -88,8 +88,6 @@ endfunction() set(CUDF_INSTALL_LIBRARY_DEPS OFF) set(CUDF_EXCLUDE_DEPS_FROM_ALL ON) set(CUDF_EXCLUDE_DEPS_FROM_ALL_FLAG EXCLUDE_FROM_ALL) -set(CUDF_ENABLE_ARROW_COMPUTE ON) -include(../cmake/thirdparty/get_arrow.cmake) # ################################################################################################## # test sources ################################################################################## @@ -223,7 +221,7 @@ ConfigureTest( QUANTILES_TEST quantiles/percentile_approx_test.cpp quantiles/quantile_test.cpp quantiles/quantiles_test.cpp GPUS 1 - PERCENT 70 EXTRA_LIBS ${ARROW_LIBRARIES} + PERCENT 70 ) # ################################################################################################## @@ -305,16 +303,13 @@ ConfigureTest( INTEROP_TEST interop/arrow_data_structures_test.cpp interop/to_arrow_device_test.cpp - interop/to_arrow_test.cpp interop/to_arrow_host_test.cpp - interop/from_arrow_test.cpp interop/from_arrow_device_test.cpp interop/from_arrow_host_test.cpp interop/from_arrow_stream_test.cpp interop/dlpack_test.cpp EXTRA_LIBS nanoarrow::nanoarrow - ${ARROW_LIBRARIES} ) # ################################################################################################## diff --git a/cpp/tests/interop/from_arrow_host_test.cpp b/cpp/tests/interop/from_arrow_host_test.cpp index 3cb451165f79..8d1ac5602fde 100644 --- a/cpp/tests/interop/from_arrow_host_test.cpp +++ b/cpp/tests/interop/from_arrow_host_test.cpp @@ -24,6 +24,92 @@ #include +#include +#include + +namespace { + +void release_schema(ArrowSchema* schema) { schema->release = nullptr; } + +void release_array(ArrowArray* array) { array->release = nullptr; } + +struct direct_arrow_c_producer { + static constexpr int64_t num_rows = 5; + + std::array int_values{1, 2, 5, 2, 7}; + std::array int_validity{0b00011101}; + + std::array string_offsets{0, 3, 6, 6, 6, 9}; + std::array string_chars{'f', 'f', 'f', 'a', 'a', 'a', 'c', 'c', 'c'}; + std::array string_validity{0b00010111}; + + ArrowSchema schema{}; + std::array child_schemas{}; + std::array child_schema_ptrs{}; + + ArrowArray array{}; + std::array child_arrays{}; + std::array child_array_ptrs{}; + std::array parent_buffers{nullptr}; + std::array int_buffers{int_validity.data(), int_values.data()}; + std::array string_buffers{ + string_validity.data(), string_offsets.data(), string_chars.data()}; + + direct_arrow_c_producer() + { + child_schema_ptrs = {&child_schemas[0], &child_schemas[1]}; + child_array_ptrs = {&child_arrays[0], &child_arrays[1]}; + + schema.format = "+s"; + schema.name = ""; + schema.flags = 0; + schema.n_children = child_schemas.size(); + schema.children = child_schema_ptrs.data(); + schema.release = release_schema; + + child_schemas[0].format = "i"; + child_schemas[0].name = "ints"; + child_schemas[0].flags = ARROW_FLAG_NULLABLE; + child_schemas[0].release = release_schema; + + child_schemas[1].format = "u"; + child_schemas[1].name = "strings"; + child_schemas[1].flags = ARROW_FLAG_NULLABLE; + child_schemas[1].release = release_schema; + + array.length = num_rows; + array.null_count = 0; + array.n_buffers = parent_buffers.size(); + array.n_children = child_arrays.size(); + array.buffers = parent_buffers.data(); + array.children = child_array_ptrs.data(); + array.release = release_array; + + child_arrays[0].length = num_rows; + child_arrays[0].null_count = 1; + child_arrays[0].n_buffers = int_buffers.size(); + child_arrays[0].buffers = int_buffers.data(); + child_arrays[0].release = release_array; + + child_arrays[1].length = num_rows; + child_arrays[1].null_count = 1; + child_arrays[1].n_buffers = string_buffers.size(); + child_arrays[1].buffers = string_buffers.data(); + child_arrays[1].release = release_array; + } + + ArrowDeviceArray device_array() const + { + ArrowDeviceArray out{}; + std::memcpy(&out.array, &array, sizeof(ArrowArray)); + out.device_type = ARROW_DEVICE_CPU; + out.device_id = -1; + return out; + } +}; + +} // namespace + // create a cudf::table and equivalent arrow table with host memory std::tuple, nanoarrow::UniqueSchema, nanoarrow::UniqueArray> get_nanoarrow_host_tables(cudf::size_type length) @@ -106,6 +192,21 @@ TEST_F(FromArrowHostDeviceTest, EmptyTable) CUDF_TEST_EXPECT_TABLES_EQUAL(expected_cudf_table, got_cudf_table->view()); } +TEST_F(FromArrowHostDeviceTest, DirectArrowCProducerTable) +{ + direct_arrow_c_producer producer; + auto input = producer.device_array(); + + auto const expected_ints = + cudf::test::fixed_width_column_wrapper{{1, 2, 5, 2, 7}, {1, 0, 1, 1, 1}}; + auto const expected_strings = + cudf::test::strings_column_wrapper{{"fff", "aaa", "", "xxx", "ccc"}, {1, 1, 1, 0, 1}}; + auto const expected = cudf::table_view{{expected_ints, expected_strings}}; + + auto got_cudf_table = cudf::from_arrow_host(&producer.schema, &input); + CUDF_TEST_EXPECT_TABLES_EQUIVALENT(expected, got_cudf_table->view()); +} + TEST_F(FromArrowHostDeviceTest, ZeroColumnsWithRows) { constexpr cudf::size_type num_rows = 5; diff --git a/cpp/tests/quantiles/percentile_approx_test.cpp b/cpp/tests/quantiles/percentile_approx_test.cpp index 301538b1e881..98e7214e3a7c 100644 --- a/cpp/tests/quantiles/percentile_approx_test.cpp +++ b/cpp/tests/quantiles/percentile_approx_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 */ @@ -14,92 +14,24 @@ #include #include #include -#include #include -#include #include #include #include #include -#include -#include -#include - namespace { -std::unique_ptr arrow_percentile_approx(cudf::column_view const& _values, - int delta, - std::vector const& percentages) -{ - static auto const _arrow_init_status = arrow::compute::Initialize(); - EXPECT_TRUE(_arrow_init_status.ok()); - - auto stream = cudf::get_default_stream(); - - // sort the incoming values using the same settings that groupby does. - // this is a little weak because null_order::AFTER is hardcoded internally to groupby. - cudf::table_view t({_values}); - auto sorted_t = cudf::sort(t, {}, {cudf::null_order::AFTER}, stream); - auto sorted_values = sorted_t->get_column(0).view(); - - std::vector h_values(sorted_values.size()); - CUDF_CUDA_TRY(cudaMemcpyAsync(h_values.data(), - sorted_values.data(), - sizeof(double) * sorted_values.size(), - cudaMemcpyDefault, - stream.value())); - std::vector h_validity(sorted_values.size()); - if (sorted_values.null_mask() != nullptr) { - auto validity = cudf::mask_to_bools(sorted_values.null_mask(), 0, sorted_values.size(), stream); - CUDF_CUDA_TRY(cudaMemcpyAsync(h_validity.data(), - (validity->view().data()), - sizeof(char) * sorted_values.size(), - cudaMemcpyDefault, - stream.value())); - } - - // generate the tdigest - arrow::DoubleBuilder builder; - for (size_t idx = 0; idx < h_values.size(); idx++) { - if (sorted_values.null_mask() == nullptr || h_validity[idx]) { - EXPECT_TRUE(builder.Append(h_values[idx]).ok()); - } - } - std::shared_ptr array; - EXPECT_TRUE(builder.Finish(&array).ok()); - - auto const udelta = static_cast(delta); - auto const usize = static_cast(h_values.size()) * 2; - arrow::compute::TDigestOptions options{percentages, udelta, usize}; - - auto arrow_result = arrow::compute::CallFunction("tdigest", {array}, &options); - auto result_array = arrow_result.ValueOrDie().array_as(); - - // copy the percentiles and stuff them into a list column - std::vector h_result; - h_result.reserve(percentages.size()); - std::transform( - result_array->begin(), result_array->end(), std::back_inserter(h_result), [](auto p) { - return p.value(); - }); - cudf::test::fixed_width_column_wrapper result(h_result.begin(), h_result.end()); - cudf::test::fixed_width_column_wrapper offsets{ - 0, static_cast(percentages.size())}; - stream.synchronize(); - return cudf::make_lists_column(1, offsets.release(), result.release(), 0, {}); -} - struct percentile_approx_dispatch { template std::unique_ptr operator()(Func op, cudf::column_view const& values, int delta, std::vector const& percentages, - cudf::size_type ulps) + [[maybe_unused]] cudf::size_type ulps) requires(cudf::is_numeric() || cudf::is_fixed_point()) { - // gpu implementation + // gpu implementation. auto agg_result = op(values, delta); cudf::test::fixed_width_column_wrapper g_percentages(percentages.begin(), @@ -107,20 +39,6 @@ struct percentile_approx_dispatch { cudf::tdigest::tdigest_column_view tdv(*agg_result); auto result = cudf::percentile_approx(tdv, g_percentages); - // disable checking logic during a racecheck run - if (getenv("LIBCUDF_RACECHECK_ENABLED")) { return result; } - - // arrow implementation. - auto expected = [&]() { - // we're explicitly casting back to doubles here but this is ok because that is - // exactly what happens inside of the cudf implementation as values are processed as well. - // so this should not affect results. - auto as_doubles = cudf::cast(values, cudf::data_type{cudf::type_id::FLOAT64}); - return arrow_percentile_approx(*as_doubles, delta, percentages); - }(); - cudf::test::detail::expect_columns_equivalent( - *expected, *result, cudf::test::debug_output_level::FIRST_ERROR, ulps); - return result; } @@ -494,6 +412,76 @@ TEST_F(PercentileApproxTest, NullPercentiles) CUDF_TEST_EXPECT_COLUMNS_EQUAL(*result, expected); } +TEST_F(PercentileApproxTest, ReductionGold) +{ + auto const delta = 1000; + + auto const values = cudf::test::fixed_width_column_wrapper{9, 1, 7, 5, 2}; + + auto const tdigest = + cudf::reduce(values, + *cudf::make_tdigest_aggregation(delta), + cudf::data_type{cudf::type_id::STRUCT}); + auto const tdigest_col = cudf::make_column_from_scalar(*tdigest, 1); + + auto const percentiles = + cudf::test::fixed_width_column_wrapper{0.0, 0.25, 0.5, 0.75, 1.0}; + auto const result = cudf::percentile_approx(tdigest_col->view(), percentiles); + + auto const expected = cudf::test::lists_column_wrapper{{1, 2, 5, 7, 9}}; + CUDF_TEST_EXPECT_COLUMNS_EQUAL(*result, expected); +} + +TEST_F(PercentileApproxTest, GroupByGold) +{ + auto const delta = 1000; + + auto const values = + cudf::test::fixed_width_column_wrapper{9, 1, 7, 5, 2, 50, 10, 40, 20, 30}; + auto const keys = cudf::test::fixed_width_column_wrapper{0, 0, 0, 0, 0, 1, 1, 1, 1, 1}; + auto const percentiles = + cudf::test::fixed_width_column_wrapper{0.0, 0.25, 0.5, 0.75, 1.0}; + + cudf::groupby::groupby gb(cudf::table_view{{keys}}); + std::vector requests; + std::vector> aggregations; + aggregations.push_back(cudf::make_tdigest_aggregation(delta)); + requests.push_back({values, std::move(aggregations)}); + auto const tdigest_column = gb.aggregate(requests); + + cudf::tdigest::tdigest_column_view tdv(*tdigest_column.second[0].results[0]); + auto const result = cudf::percentile_approx(tdv, percentiles); + + auto const expected = + cudf::test::lists_column_wrapper{{1, 2, 5, 7, 9}, {10, 20, 30, 40, 50}}; + CUDF_TEST_EXPECT_COLUMNS_EQUAL(*result, expected); +} + +TEST_F(PercentileApproxTest, GroupByWithNullsGold) +{ + auto const delta = 1000; + + auto const values = cudf::test::fixed_width_column_wrapper{ + {9, 99, 7, 5, 1, 50, 10, 40, 20, 30}, {1, 0, 1, 1, 1, 1, 1, 1, 1, 1}}; + auto const keys = cudf::test::fixed_width_column_wrapper{0, 0, 0, 0, 0, 1, 1, 1, 1, 1}; + auto const percentiles = + cudf::test::fixed_width_column_wrapper{0.0, 0.25, 0.5, 0.75, 1.0}; + + cudf::groupby::groupby gb(cudf::table_view{{keys}}); + std::vector requests; + std::vector> aggregations; + aggregations.push_back(cudf::make_tdigest_aggregation(delta)); + requests.push_back({values, std::move(aggregations)}); + auto const tdigest_column = gb.aggregate(requests); + + cudf::tdigest::tdigest_column_view tdv(*tdigest_column.second[0].results[0]); + auto const result = cudf::percentile_approx(tdv, percentiles); + + auto const expected = + cudf::test::lists_column_wrapper{{1, 1, 5, 7, 9}, {10, 20, 30, 40, 50}}; + CUDF_TEST_EXPECT_COLUMNS_EQUAL(*result, expected); +} + TEST_F(PercentileApproxTest, ReductionWithLowRowCount) { // Test that the tdigest reduction with a low row count still produces the correct results.