Add test

Signed-off-by: Nghia Truong <[email protected]>

Author: ttnghia

cpp/tests/CMakeLists.txt

@@ -122,44 +122,7 @@ ConfigureTest(TIMESTAMPS_TEST wrappers/timestamps_test.cu)
 # * groupby tests ---------------------------------------------------------------------------------
 ConfigureTest(
   GROUPBY_TEST
-  groupby/argmin_tests.cpp
-  groupby/argmax_tests.cpp
-  groupby/collect_list_tests.cpp
-  groupby/collect_set_tests.cpp
-  groupby/correlation_tests.cpp
-  groupby/count_scan_tests.cpp
-  groupby/count_tests.cpp
-  groupby/covariance_tests.cpp
-  groupby/groupby_test_util.cpp
-  groupby/groups_tests.cpp
-  groupby/histogram_tests.cpp
-  groupby/keys_tests.cpp
-  groupby/lists_tests.cpp
-  groupby/m2_tests.cpp
-  groupby/min_tests.cpp
-  groupby/max_scan_tests.cpp
-  groupby/max_tests.cpp
-  groupby/mean_tests.cpp
-  groupby/median_tests.cpp
-  groupby/merge_m2_tests.cpp
-  groupby/merge_lists_tests.cpp
-  groupby/merge_sets_tests.cpp
-  groupby/min_scan_tests.cpp
-  groupby/nth_element_tests.cpp
-  groupby/nunique_tests.cpp
-  groupby/product_scan_tests.cpp
-  groupby/product_tests.cpp
-  groupby/quantile_tests.cpp
-  groupby/rank_scan_tests.cpp
-  groupby/replace_nulls_tests.cpp
-  groupby/shift_tests.cpp
-  groupby/std_tests.cpp
-  groupby/structs_tests.cpp
-  groupby/sum_of_squares_tests.cpp
-  groupby/sum_scan_tests.cpp
-  groupby/sum_tests.cpp
-  groupby/tdigest_tests.cu
-  groupby/var_tests.cpp
+  groupby/host_udf_tests.cu
   GPUS 1
   PERCENT 100
 )

cpp/tests/groupby/host_udf_tests.cu

@@ -0,0 +1,111 @@
+/*
+ * Copyright (c) 2019-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 <tests/groupby/groupby_test_util.hpp>
+
+#include <cudf_test/base_fixture.hpp>
+#include <cudf_test/column_wrapper.hpp>
+#include <cudf_test/debug_utilities.hpp>
+#include <cudf_test/iterator_utilities.hpp>
+#include <cudf_test/type_lists.hpp>
+
+#include <cudf/column/column_factories.hpp>
+#include <cudf/detail/aggregation/aggregation.hpp>
+
+#include <rmm/exec_policy.hpp>
+
+#include <thrust/iterator/counting_iterator.h>
+#include <thrust/transform.h>
+
+using namespace cudf::test::iterators;
+
+struct test : public cudf::test::BaseFixture {};
+
+std::unique_ptr<cudf::column> double_sqr(cudf::column_view const& values,
+                                         cudf::device_span<cudf::size_type const> group_offsets,
+                                         cudf::device_span<cudf::size_type const> group_labels,
+                                         cudf::size_type num_groups,
+                                         rmm::cuda_stream_view stream,
+                                         rmm::device_async_resource_ref mr)
+{
+  auto output = cudf::make_numeric_column(
+    cudf::data_type{cudf::type_id::INT32}, values.size(), cudf::mask_state::UNALLOCATED, stream);
+  thrust::transform(rmm::exec_policy(stream),
+                    thrust::make_counting_iterator(0),
+                    thrust::make_counting_iterator(values.size()),
+                    output->mutable_view().begin<int>(),
+                    [values = values.begin<int>()] __device__(int idx) -> int {
+                      return 2 * values[idx] * values[idx];
+                    });
+  return output;
+}
+
+std::unique_ptr<cudf::column> triple_sqr(cudf::column_view const& values,
+                                         cudf::device_span<cudf::size_type const> group_offsets,
+                                         cudf::device_span<cudf::size_type const> group_labels,
+                                         cudf::size_type num_groups,
+                                         rmm::cuda_stream_view stream,
+                                         rmm::device_async_resource_ref mr)
+{
+  auto output = cudf::make_numeric_column(
+    cudf::data_type{cudf::type_id::INT32}, values.size(), cudf::mask_state::UNALLOCATED, stream);
+  thrust::transform(rmm::exec_policy(stream),
+                    thrust::make_counting_iterator(0),
+                    thrust::make_counting_iterator(values.size()),
+                    output->mutable_view().begin<int>(),
+                    [values = values.begin<int>()] __device__(int idx) -> int {
+                      return 3 * values[idx] * values[idx];
+                    });
+  return output;
+}
+
+TEST_F(test, double_sqr)
+{
+  cudf::test::fixed_width_column_wrapper<int> keys{1, 1, 1, 1, 1};
+  cudf::test::fixed_width_column_wrapper<int> vals{0, 1, 2, 3, 4};
+
+  auto agg = cudf::make_host_udf_aggregation<cudf::groupby_aggregation>(double_sqr);
+  std::vector<cudf::groupby::aggregation_request> requests;
+  requests.emplace_back();
+  requests[0].values = vals;
+  requests[0].aggregations.push_back(std::move(agg));
+  cudf::groupby::groupby gb_obj(
+    cudf::table_view({keys}), cudf::null_policy::INCLUDE, cudf::sorted::NO, {}, {});
+
+  auto result = gb_obj.aggregate(requests, cudf::test::get_default_stream());
+
+  // Got output: 0,2,8,18,32
+  cudf::test::print(*result.second[0].results[0]);
+}
+
+TEST_F(test, triple_sqr)
+{
+  cudf::test::fixed_width_column_wrapper<int> keys{1, 1, 1, 1, 1};
+  cudf::test::fixed_width_column_wrapper<int> vals{0, 1, 2, 3, 4};
+
+  auto agg = cudf::make_host_udf_aggregation<cudf::groupby_aggregation>(triple_sqr);
+  std::vector<cudf::groupby::aggregation_request> requests;
+  requests.emplace_back();
+  requests[0].values = vals;
+  requests[0].aggregations.push_back(std::move(agg));
+  cudf::groupby::groupby gb_obj(
+    cudf::table_view({keys}), cudf::null_policy::INCLUDE, cudf::sorted::NO, {}, {});
+
+  auto result = gb_obj.aggregate(requests, cudf::test::get_default_stream());
+
+  // Got output: 0,3,12,27,48
+  cudf::test::print(*result.second[0].results[0]);
+}