diff --git a/cpp/include/cudf_test/base_fixture.hpp b/cpp/include/cudf_test/base_fixture.hpp index 583abe9931d8..003955369001 100644 --- a/cpp/include/cudf_test/base_fixture.hpp +++ b/cpp/include/cudf_test/base_fixture.hpp @@ -1,17 +1,20 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once #include +#include #include +#include #include #include #include +#include #include #include @@ -39,6 +42,34 @@ class BaseFixture : public ::testing::Test { rmm::device_async_resource_ref mr() { return _mr; } }; +/** + * @brief Base fixture that instruments tests with a memory-resource harness. + * + * Each test instantiates a fresh harness. Tests should construct results with `resources()`. + * `TearDown` asserts that no output or temporary allocations remain live. + */ +struct BaseFixtureWithHarness : public BaseFixture { + /** + * @brief Assert that the harness has no live output or temporary allocations. + */ + void TearDown() override { _harness.expect_no_live_allocations(stream()); } + + /** + * @brief Return the default stream used by tests inheriting from this fixture. + * @return CUDA stream view + */ + [[nodiscard]] rmm::cuda_stream_view stream() const { return cudf::test::get_default_stream(); } + + /** + * @brief Return the harness output and temporary memory resources. + * @return Explicit output and temporary resources that do not consult the current resource + */ + cudf::memory_resources resources() { return _harness.resources(); } + + protected: + memory_resource_test_harness _harness{mr()}; +}; + /** * @brief Base test fixture that takes a parameter. * diff --git a/cpp/include/cudf_test/column_wrapper.hpp b/cpp/include/cudf_test/column_wrapper.hpp index 632a24155cf2..54acc5c5e7a9 100644 --- a/cpp/include/cudf_test/column_wrapper.hpp +++ b/cpp/include/cudf_test/column_wrapper.hpp @@ -26,6 +26,7 @@ #include #include +#include #include #include @@ -147,21 +148,27 @@ struct fixed_width_type_converter { * @tparam InputIterator Iterator type for `begin` and `end` * @param begin Beginning of the sequence of elements * @param end End of the sequence of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned buffer * @return rmm::device_buffer Buffer containing all elements in the range `[begin,end)` */ template ()>* = nullptr> -rmm::device_buffer make_elements(InputIterator begin, InputIterator end) +rmm::device_buffer make_elements(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { static_assert(cudf::is_fixed_width(), "Unexpected non-fixed width type."); auto transformer = fixed_width_type_converter{}; auto transform_begin = thrust::make_transform_iterator(begin, transformer); auto const size = cudf::distance(begin, end); auto const elements = thrust::host_vector(transform_begin, transform_begin + size); - return rmm::device_buffer{ - elements.data(), size * sizeof(ElementTo), cudf::test::get_default_stream()}; + rmm::device_buffer buffer{elements.data(), size * sizeof(ElementTo), stream, mr.get_output_mr()}; + stream.synchronize(); // wait for async H2D before host source is destroyed + return buffer; } // The two signatures below are identical to the above overload apart from @@ -176,6 +183,8 @@ rmm::device_buffer make_elements(InputIterator begin, InputIterator end) * @tparam InputIterator Iterator type for `begin` and `end` * @param begin Beginning of the sequence of elements * @param end End of the sequence of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned buffer * @return rmm::device_buffer Buffer containing all elements in the range `[begin,end)` */ template () and cudf::is_fixed_point()>* = nullptr> -rmm::device_buffer make_elements(InputIterator begin, InputIterator end) +rmm::device_buffer make_elements(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { using RepType = typename ElementTo::rep; auto transformer = fixed_width_type_converter{}; auto transform_begin = thrust::make_transform_iterator(begin, transformer); auto const size = cudf::distance(begin, end); auto const elements = thrust::host_vector(transform_begin, transform_begin + size); - return rmm::device_buffer{ - elements.data(), size * sizeof(RepType), cudf::test::get_default_stream()}; + rmm::device_buffer buffer{elements.data(), size * sizeof(RepType), stream, mr.get_output_mr()}; + stream.synchronize(); // wait for async H2D before host source is destroyed + return buffer; } /** @@ -202,6 +215,8 @@ rmm::device_buffer make_elements(InputIterator begin, InputIterator end) * @tparam InputIterator Iterator type for `begin` and `end` * @param begin Beginning of the sequence of elements * @param end End of the sequence of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned buffer * @return rmm::device_buffer Buffer containing all elements in the range `[begin,end)` */ template () and cudf::is_fixed_point()>* = nullptr> -rmm::device_buffer make_elements(InputIterator begin, InputIterator end) +rmm::device_buffer make_elements(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { using namespace numeric; using RepType = typename ElementTo::rep; @@ -221,8 +239,9 @@ rmm::device_buffer make_elements(InputIterator begin, InputIterator end) auto transformer_begin = thrust::make_transform_iterator(begin, to_rep); auto const size = cudf::distance(begin, end); auto const elements = thrust::host_vector(transformer_begin, transformer_begin + size); - return rmm::device_buffer{ - elements.data(), size * sizeof(RepType), cudf::test::get_default_stream()}; + rmm::device_buffer buffer{elements.data(), size * sizeof(RepType), stream, mr.get_output_mr()}; + stream.synchronize(); // wait for async H2D before host source is destroyed + return buffer; } //! @endcond @@ -269,17 +288,24 @@ std::pair, cudf::size_type> make_null_mask_vector(Vali * @tparam ValidityIterator * @param begin The beginning of the validity indicator sequence * @param end The end of the validity indicator sequence + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned buffer * @return rmm::device_buffer Contains a bitmask where bits are set for every * element in `[begin,end)` that evaluated to `true`. */ template -std::pair make_null_mask(ValidityIterator begin, - ValidityIterator end) +std::pair make_null_mask( + ValidityIterator begin, + ValidityIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { auto [null_mask, null_count] = make_null_mask_vector(begin, end); - auto d_mask = rmm::device_buffer{null_mask.data(), - cudf::bitmask_allocation_size_bytes(cudf::distance(begin, end)), - cudf::test::get_default_stream()}; + rmm::device_buffer d_mask{null_mask.data(), + cudf::bitmask_allocation_size_bytes(cudf::distance(begin, end)), + stream, + mr.get_output_mr()}; + stream.synchronize(); // wait for async H2D before host source is destroyed return {std::move(d_mask), null_count}; } @@ -332,13 +358,7 @@ class fixed_width_column_wrapper : public detail::column_wrapper { */ fixed_width_column_wrapper() : column_wrapper{} { - std::vector empty; - wrapped.reset( - new cudf::column{cudf::data_type{cudf::type_to_id()}, - 0, - detail::make_elements(empty.begin(), empty.end()), - rmm::device_buffer{}, - 0}); + wrapped = cudf::make_empty_column(cudf::type_to_id()); } /** @@ -358,16 +378,23 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - fixed_width_column_wrapper(InputIterator begin, InputIterator end) : column_wrapper{} + fixed_width_column_wrapper(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { auto const size = cudf::distance(begin, end); - wrapped.reset(new cudf::column{cudf::data_type{cudf::type_to_id()}, - size, - detail::make_elements(begin, end), - rmm::device_buffer{}, - 0}); + wrapped.reset( + new cudf::column{cudf::data_type{cudf::type_to_id()}, + size, + detail::make_elements(begin, end, stream, mr), + rmm::device_buffer{}, + 0}); } /** @@ -392,18 +419,25 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - fixed_width_column_wrapper(InputIterator begin, InputIterator end, ValidityIterator v) + fixed_width_column_wrapper(InputIterator begin, + InputIterator end, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { auto const size = cudf::distance(begin, end); - auto [null_mask, null_count] = detail::make_null_mask(v, v + size); - wrapped.reset(new cudf::column{cudf::data_type{cudf::type_to_id()}, - size, - detail::make_elements(begin, end), - std::move(null_mask), - null_count}); + auto [null_mask, null_count] = detail::make_null_mask(v, v + size, stream, mr); + wrapped.reset( + new cudf::column{cudf::data_type{cudf::type_to_id()}, + size, + detail::make_elements(begin, end, stream, mr), + std::move(null_mask), + null_count}); } /** @@ -417,10 +451,14 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - fixed_width_column_wrapper(std::initializer_list elements) - : fixed_width_column_wrapper(std::cbegin(elements), std::cend(elements)) + fixed_width_column_wrapper(std::initializer_list elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_width_column_wrapper(std::cbegin(elements), std::cend(elements), stream, mr) { } @@ -440,11 +478,16 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * * @param elements The list of elements * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_width_column_wrapper(std::initializer_list elements, - std::initializer_list validity) - : fixed_width_column_wrapper(std::cbegin(elements), std::cend(elements), std::cbegin(validity)) + std::initializer_list validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_width_column_wrapper( + std::cbegin(elements), std::cend(elements), std::cbegin(validity), stream, mr) { } @@ -464,10 +507,15 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * convertible to `bool` * @param element_list The list of elements * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - fixed_width_column_wrapper(std::initializer_list element_list, ValidityIterator v) - : fixed_width_column_wrapper(std::cbegin(element_list), std::cend(element_list), v) + fixed_width_column_wrapper(std::initializer_list element_list, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_width_column_wrapper(std::cbegin(element_list), std::cend(element_list), v, stream, mr) { } @@ -488,12 +536,16 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_width_column_wrapper(InputIterator begin, InputIterator end, - std::initializer_list const& validity) - : fixed_width_column_wrapper(begin, end, std::cbegin(validity)) + std::initializer_list const& validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_width_column_wrapper(begin, end, std::cbegin(validity), stream, mr) { } @@ -513,16 +565,21 @@ class fixed_width_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of pairs of element and validity booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - fixed_width_column_wrapper(std::initializer_list> elements) + fixed_width_column_wrapper(std::initializer_list> elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { auto begin = thrust::make_transform_iterator(elements.begin(), [](auto const& e) { return e.first; }); auto end = begin + elements.size(); auto v = thrust::make_transform_iterator(elements.begin(), [](auto const& e) { return e.second; }); - wrapped = fixed_width_column_wrapper(begin, end, v).release(); + wrapped = + fixed_width_column_wrapper(begin, end, v, stream, mr).release(); } }; @@ -549,11 +606,15 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_point_column_wrapper(FixedPointRepIterator begin, FixedPointRepIterator end, - numeric::scale_type scale) + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { CUDF_EXPECTS(numeric::is_supported_representation_type(), "not valid representation type"); @@ -562,13 +623,9 @@ class fixed_point_column_wrapper : public detail::column_wrapper { auto const elements = thrust::host_vector(begin, end); auto const id = type_to_id>(); auto const data_type = cudf::data_type{id, static_cast(scale)}; - - wrapped.reset(new cudf::column{ - data_type, - size, - rmm::device_buffer{elements.data(), size * sizeof(Rep), cudf::test::get_default_stream()}, - rmm::device_buffer{}, - 0}); + rmm::device_buffer data{elements.data(), size * sizeof(Rep), stream, mr.get_output_mr()}; + wrapped.reset(new cudf::column{data_type, size, std::move(data), rmm::device_buffer{}, 0}); + stream.synchronize(); // wait for async H2D before host source is destroyed } /** @@ -582,9 +639,14 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * * @param values The initializer list of already shifted values * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ - fixed_point_column_wrapper(std::initializer_list values, numeric::scale_type scale) - : fixed_point_column_wrapper(std::cbegin(values), std::cend(values), scale) + fixed_point_column_wrapper(std::initializer_list values, + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_point_column_wrapper(std::cbegin(values), std::cend(values), scale, stream, mr) { } @@ -614,12 +676,16 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * @param end The end of the sequence of elements * @param v The beginning of the sequence of validity indicators * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_point_column_wrapper(FixedPointRepIterator begin, FixedPointRepIterator end, ValidityIterator v, - numeric::scale_type scale) + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { CUDF_EXPECTS(numeric::is_supported_representation_type(), "not valid representation type"); @@ -628,13 +694,11 @@ class fixed_point_column_wrapper : public detail::column_wrapper { auto const elements = thrust::host_vector(begin, end); auto const id = type_to_id>(); auto const data_type = cudf::data_type{id, static_cast(scale)}; - auto [null_mask, null_count] = detail::make_null_mask(v, v + size); - wrapped.reset(new cudf::column{ - data_type, - size, - rmm::device_buffer{elements.data(), size * sizeof(Rep), cudf::test::get_default_stream()}, - std::move(null_mask), - null_count}); + auto [null_mask, null_count] = detail::make_null_mask(v, v + size, stream, mr); + rmm::device_buffer data{elements.data(), size * sizeof(Rep), stream, mr.get_output_mr()}; + wrapped.reset( + new cudf::column{data_type, size, std::move(data), std::move(null_mask), null_count}); + stream.synchronize(); // wait for async H2D before host source is destroyed } /** @@ -653,12 +717,16 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * @param elements The initializer list of elements * @param validity The initializer list of validity indicator booleans * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ fixed_point_column_wrapper(std::initializer_list elements, std::initializer_list validity, - numeric::scale_type scale) + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : fixed_point_column_wrapper( - std::cbegin(elements), std::cend(elements), std::cbegin(validity), scale) + std::cbegin(elements), std::cend(elements), std::cbegin(validity), scale, stream, mr) { } @@ -679,12 +747,17 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * @param element_list The initializer list of elements * @param v The beginning of the sequence of validity indicators * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_point_column_wrapper(std::initializer_list element_list, ValidityIterator v, - numeric::scale_type scale) - : fixed_point_column_wrapper(std::cbegin(element_list), std::cend(element_list), v, scale) + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_point_column_wrapper( + std::cbegin(element_list), std::cend(element_list), v, scale, stream, mr) { } @@ -707,13 +780,17 @@ class fixed_point_column_wrapper : public detail::column_wrapper { * @param end The end of the sequence of elements * @param validity The initializer list of validity indicator booleans * @param scale The scale of the elements in the column + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template fixed_point_column_wrapper(FixedPointRepIterator begin, FixedPointRepIterator end, std::initializer_list const& validity, - numeric::scale_type scale) - : fixed_point_column_wrapper(begin, end, std::cbegin(validity), scale) + numeric::scale_type scale, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : fixed_point_column_wrapper(begin, end, std::cbegin(validity), scale, stream, mr) { } }; @@ -726,7 +803,10 @@ class strings_column_wrapper : public detail::column_wrapper { /** * @brief Default constructor initializes an empty column of strings */ - strings_column_wrapper() : strings_column_wrapper(std::initializer_list{}) {} + strings_column_wrapper() : column_wrapper{} + { + wrapped = cudf::make_empty_column(cudf::type_id::STRING); + } /** * @brief Construct a non-nullable column of strings from the range @@ -747,9 +827,15 @@ class strings_column_wrapper : public detail::column_wrapper { * dereferencing a `StringsIterator`. * @param begin The beginning of the sequence * @param end The end of the sequence + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - strings_column_wrapper(StringsIterator begin, StringsIterator end) : column_wrapper{} + strings_column_wrapper(StringsIterator begin, + StringsIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { size_type num_strings = std::distance(begin, end); if (num_strings == 0) { @@ -758,11 +844,9 @@ class strings_column_wrapper : public detail::column_wrapper { } auto all_valid = cuda::make_constant_iterator(true); auto [chars, offsets] = detail::make_chars_and_offsets(begin, end, all_valid); - auto d_chars = cudf::detail::make_device_uvector_async( - chars, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref()); + auto d_chars = cudf::detail::make_device_uvector_async(chars, stream, mr.get_output_mr()); auto d_offsets = std::make_unique( - cudf::detail::make_device_uvector( - offsets, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref()), + cudf::detail::make_device_uvector(offsets, stream, mr.get_output_mr()), rmm::device_buffer{}, 0); wrapped = @@ -796,9 +880,15 @@ class strings_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence * @param end The end of the sequence * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - strings_column_wrapper(StringsIterator begin, StringsIterator end, ValidityIterator v) + strings_column_wrapper(StringsIterator begin, + StringsIterator end, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { size_type num_strings = std::distance(begin, end); @@ -808,16 +898,13 @@ class strings_column_wrapper : public detail::column_wrapper { } auto [chars, offsets] = detail::make_chars_and_offsets(begin, end, v); auto [null_mask, null_count] = detail::make_null_mask_vector(v, v + num_strings); - auto d_chars = cudf::detail::make_device_uvector_async( - chars, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref()); + auto d_chars = cudf::detail::make_device_uvector_async(chars, stream, mr.get_output_mr()); auto d_offsets = std::make_unique( - cudf::detail::make_device_uvector_async( - offsets, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref()), + cudf::detail::make_device_uvector_async(offsets, stream, mr.get_output_mr()), rmm::device_buffer{}, 0); - auto d_bitmask = cudf::detail::make_device_uvector( - null_mask, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref()); - wrapped = cudf::make_strings_column( + auto d_bitmask = cudf::detail::make_device_uvector(null_mask, stream, mr.get_output_mr()); + wrapped = cudf::make_strings_column( num_strings, std::move(d_offsets), d_chars.release(), null_count, d_bitmask.release()); } @@ -832,9 +919,13 @@ class strings_column_wrapper : public detail::column_wrapper { * @endcode * * @param strings The list of strings + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ - strings_column_wrapper(std::initializer_list strings) - : strings_column_wrapper(std::cbegin(strings), std::cend(strings)) + strings_column_wrapper(std::initializer_list strings, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : strings_column_wrapper(std::cbegin(strings), std::cend(strings), stream, mr) { } @@ -855,10 +946,15 @@ class strings_column_wrapper : public detail::column_wrapper { * convertible to `bool` * @param strings The list of strings * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - strings_column_wrapper(std::initializer_list strings, ValidityIterator v) - : strings_column_wrapper(std::cbegin(strings), std::cend(strings), v) + strings_column_wrapper(std::initializer_list strings, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : strings_column_wrapper(std::cbegin(strings), std::cend(strings), v, stream, mr) { } @@ -876,10 +972,15 @@ class strings_column_wrapper : public detail::column_wrapper { * * @param strings The list of strings * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ strings_column_wrapper(std::initializer_list strings, - std::initializer_list validity) - : strings_column_wrapper(std::cbegin(strings), std::cend(strings), std::cbegin(validity)) + std::initializer_list validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : strings_column_wrapper( + std::cbegin(strings), std::cend(strings), std::cbegin(validity), stream, mr) { } @@ -902,15 +1003,19 @@ class strings_column_wrapper : public detail::column_wrapper { * @endcode * * @param strings The list of pairs of strings and validity booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ - strings_column_wrapper(std::initializer_list> strings) + strings_column_wrapper(std::initializer_list> strings, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { auto begin = thrust::make_transform_iterator(strings.begin(), [](auto const& s) { return s.first; }); auto end = begin + strings.size(); auto v = thrust::make_transform_iterator(strings.begin(), [](auto const& s) { return s.second; }); - wrapped = strings_column_wrapper(begin, end, v).release(); + wrapped = strings_column_wrapper(begin, end, v, stream, mr).release(); } }; @@ -957,15 +1062,21 @@ class dictionary_column_wrapper : public detail::column_wrapper { * * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(InputIterator begin, InputIterator end) : column_wrapper{} + dictionary_column_wrapper(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { - wrapped = - cudf::dictionary::encode(fixed_width_column_wrapper(begin, end), - cudf::data_type{type_id::INT32}, - cudf::test::get_default_stream(), - cudf::get_current_device_resource_ref()); + wrapped = cudf::dictionary::encode(fixed_width_column_wrapper( + begin, end, stream, mr.get_temporary_mr()), + cudf::data_type{type_id::INT32}, + stream, + mr.get_output_mr()); } /** @@ -992,15 +1103,22 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(InputIterator begin, InputIterator end, ValidityIterator v) + dictionary_column_wrapper(InputIterator begin, + InputIterator end, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { - wrapped = cudf::dictionary::encode( - fixed_width_column_wrapper(begin, end, v), - cudf::data_type{type_id::INT32}, - cudf::test::get_default_stream()); + wrapped = cudf::dictionary::encode(fixed_width_column_wrapper( + begin, end, v, stream, mr.get_temporary_mr()), + cudf::data_type{type_id::INT32}, + stream, + mr.get_output_mr()); } /** @@ -1015,10 +1133,14 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(std::initializer_list elements) - : dictionary_column_wrapper(std::cbegin(elements), std::cend(elements)) + dictionary_column_wrapper(std::initializer_list elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper(std::cbegin(elements), std::cend(elements), stream, mr) { } @@ -1039,11 +1161,16 @@ class dictionary_column_wrapper : public detail::column_wrapper { * * @param elements The list of elements * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template dictionary_column_wrapper(std::initializer_list elements, - std::initializer_list validity) - : dictionary_column_wrapper(std::cbegin(elements), std::cend(elements), std::cbegin(validity)) + std::initializer_list validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper( + std::cbegin(elements), std::cend(elements), std::cbegin(validity), stream, mr) { } @@ -1064,10 +1191,15 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @tparam ValidityIterator Dereferencing a ValidityIterator must be convertible to `bool` * @param element_list The list of elements * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(std::initializer_list element_list, ValidityIterator v) - : dictionary_column_wrapper(std::cbegin(element_list), std::cend(element_list), v) + dictionary_column_wrapper(std::initializer_list element_list, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper(std::cbegin(element_list), std::cend(element_list), v, stream, mr) { } @@ -1090,12 +1222,16 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence of elements * @param end The end of the sequence of elements * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template dictionary_column_wrapper(InputIterator begin, InputIterator end, - std::initializer_list const& validity) - : dictionary_column_wrapper(begin, end, std::cbegin(validity)) + std::initializer_list const& validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper(begin, end, std::cbegin(validity), stream, mr) { } }; @@ -1158,14 +1294,21 @@ class dictionary_column_wrapper : public detail::column_wrapper { * dereferencing a `StringsIterator`. * @param begin The beginning of the sequence * @param end The end of the sequence + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(StringsIterator begin, StringsIterator end) : column_wrapper{} + dictionary_column_wrapper(StringsIterator begin, + StringsIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { - wrapped = cudf::dictionary::encode(strings_column_wrapper(begin, end), - cudf::data_type{type_id::INT32}, - cudf::test::get_default_stream(), - cudf::get_current_device_resource_ref()); + wrapped = + cudf::dictionary::encode(strings_column_wrapper(begin, end, stream, mr.get_temporary_mr()), + cudf::data_type{type_id::INT32}, + stream, + mr.get_output_mr()); } /** @@ -1195,14 +1338,22 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @param begin The beginning of the sequence * @param end The end of the sequence * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(StringsIterator begin, StringsIterator end, ValidityIterator v) + dictionary_column_wrapper(StringsIterator begin, + StringsIterator end, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { - wrapped = cudf::dictionary::encode(strings_column_wrapper(begin, end, v), - cudf::data_type{type_id::INT32}, - cudf::test::get_default_stream()); + wrapped = + cudf::dictionary::encode(strings_column_wrapper(begin, end, v, stream, mr.get_temporary_mr()), + cudf::data_type{type_id::INT32}, + stream, + mr.get_output_mr()); } /** @@ -1216,9 +1367,13 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @endcode * * @param strings The list of strings + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ - dictionary_column_wrapper(std::initializer_list strings) - : dictionary_column_wrapper(std::cbegin(strings), std::cend(strings)) + dictionary_column_wrapper(std::initializer_list strings, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper(std::cbegin(strings), std::cend(strings), stream, mr) { } @@ -1239,10 +1394,15 @@ class dictionary_column_wrapper : public detail::column_wrapper { * @tparam ValidityIterator Dereferencing a ValidityIterator must be convertible to `bool` * @param strings The list of strings * @param v The beginning of the sequence of validity indicators + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template - dictionary_column_wrapper(std::initializer_list strings, ValidityIterator v) - : dictionary_column_wrapper(std::cbegin(strings), std::cend(strings), v) + dictionary_column_wrapper(std::initializer_list strings, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper(std::cbegin(strings), std::cend(strings), v, stream, mr) { } @@ -1260,10 +1420,15 @@ class dictionary_column_wrapper : public detail::column_wrapper { * * @param strings The list of strings * @param validity The list of validity indicator booleans + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ dictionary_column_wrapper(std::initializer_list strings, - std::initializer_list validity) - : dictionary_column_wrapper(std::cbegin(strings), std::cend(strings), std::cbegin(validity)) + std::initializer_list validity, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : dictionary_column_wrapper( + std::cbegin(strings), std::cend(strings), std::cbegin(validity), stream, mr) { } }; @@ -1323,12 +1488,19 @@ class lists_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template ()>* = nullptr> - lists_column_wrapper(std::initializer_list elements) : column_wrapper{} + lists_column_wrapper(std::initializer_list elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { build_from_non_nested( - cudf::test::fixed_width_column_wrapper(elements).release()); + cudf::test::fixed_width_column_wrapper(elements, stream, mr).release(), + stream, + mr); } /** @@ -1345,14 +1517,22 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param begin Beginning of the sequence * @param end End of the sequence + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template ()>* = nullptr> - lists_column_wrapper(InputIterator begin, InputIterator end) : column_wrapper{} + lists_column_wrapper(InputIterator begin, + InputIterator end, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { build_from_non_nested( - cudf::test::fixed_width_column_wrapper(begin, end).release()); + cudf::test::fixed_width_column_wrapper(begin, end, stream, mr).release(), + stream, + mr); } /** @@ -1369,15 +1549,22 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param elements The list of elements * @param v The validity iterator + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template ()>* = nullptr> - lists_column_wrapper(std::initializer_list elements, ValidityIterator v) + lists_column_wrapper(std::initializer_list elements, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { build_from_non_nested( - cudf::test::fixed_width_column_wrapper(elements, v).release()); + cudf::test::fixed_width_column_wrapper(elements, v, stream, mr).release(), + stream, + mr); } /** @@ -1396,16 +1583,25 @@ class lists_column_wrapper : public detail::column_wrapper { * @param begin Beginning of the sequence * @param end End of the sequence * @param v The validity iterator + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template ()>* = nullptr> - lists_column_wrapper(InputIterator begin, InputIterator end, ValidityIterator v) + lists_column_wrapper(InputIterator begin, + InputIterator end, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { build_from_non_nested( - cudf::test::fixed_width_column_wrapper(begin, end, v).release()); + cudf::test::fixed_width_column_wrapper(begin, end, v, stream, mr) + .release(), + stream, + mr); } /** @@ -1420,13 +1616,20 @@ class lists_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template >* = nullptr> - lists_column_wrapper(std::initializer_list elements) : column_wrapper{} + lists_column_wrapper(std::initializer_list elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + : column_wrapper{} { build_from_non_nested( - cudf::test::strings_column_wrapper(elements.begin(), elements.end()).release()); + cudf::test::strings_column_wrapper(elements.begin(), elements.end(), stream, mr).release(), + stream, + mr); } /** @@ -1443,15 +1646,22 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param elements The list of elements * @param v The validity iterator + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template >* = nullptr> - lists_column_wrapper(std::initializer_list elements, ValidityIterator v) + lists_column_wrapper(std::initializer_list elements, + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { build_from_non_nested( - cudf::test::strings_column_wrapper(elements.begin(), elements.end(), v).release()); + cudf::test::strings_column_wrapper(elements.begin(), elements.end(), v, stream, mr).release(), + stream, + mr); } /** @@ -1474,16 +1684,20 @@ class lists_column_wrapper : public detail::column_wrapper { * @endcode * * @param elements The list of elements + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ - lists_column_wrapper(std::initializer_list> elements) + lists_column_wrapper(std::initializer_list> elements, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { std::vector valids; - build_from_nested(elements, valids); + build_from_nested(elements, valids, stream, mr); } /** - * @brief Construct am empty lists column + * @brief Construct an empty lists column * * Example: * @code{.cpp} @@ -1491,11 +1705,14 @@ class lists_column_wrapper : public detail::column_wrapper { * // [] * lists_column_wrapper l{}; * @endcode - * */ lists_column_wrapper() : column_wrapper{} { - build_from_non_nested(make_empty_column(cudf::type_to_id())); + // Mark as a root so nesting unwraps to the empty child, matching + // build_from_non_nested on an empty leaf. + root = true; + depth = 0; + wrapped = make_empty_lists_column(data_type{type_to_id()}); } /** @@ -1522,10 +1739,14 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param elements The list of elements * @param v The validity iterator + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template lists_column_wrapper(std::initializer_list> elements, - ValidityIterator v) + ValidityIterator v, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) : column_wrapper{} { std::vector validity; @@ -1534,25 +1755,31 @@ class lists_column_wrapper : public detail::column_wrapper { v, std::back_inserter(validity), [](lists_column_wrapper const& l, bool valid) { return valid; }); - build_from_nested(elements, validity); + build_from_nested(elements, validity, stream, mr); } /** * @brief Construct a list column containing a single empty, optionally null row. * * @param valid Whether or not the empty row is also null + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column * @return A list column containing a single empty row */ - static lists_column_wrapper make_one_empty_row_column(bool valid = true) + static lists_column_wrapper make_one_empty_row_column( + bool valid = true, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { - cudf::test::fixed_width_column_wrapper offsets{0, 0}; + cudf::test::fixed_width_column_wrapper offsets({0, 0}, stream, mr); cudf::test::fixed_width_column_wrapper values{}; return lists_column_wrapper( 1, offsets.release(), values.release(), valid ? 0 : 1, - valid ? rmm::device_buffer{} : cudf::create_null_mask(1, cudf::mask_state::ALL_NULL)); + valid ? rmm::device_buffer{} + : cudf::create_null_mask(1, cudf::mask_state::ALL_NULL, stream, mr.get_output_mr())); } private: @@ -1590,10 +1817,14 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param elements Input columns to be wrapped * @param v The validity of each row + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column * */ void build_from_nested(std::initializer_list> elements, - std::vector const& v) + std::vector const& v, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { auto valids = cudf::detail::make_counting_transform_iterator( 0, [&v](auto i) { return v.empty() ? true : v[i]; }); @@ -1610,7 +1841,8 @@ class lists_column_wrapper : public detail::column_wrapper { int32_t const expected_depth = hierarchy_and_depth.second; // preprocess columns so that every column_view in 'cols' is an equivalent hierarchy - auto [cols, stubs] = preprocess_columns(elements, expected_hierarchy, expected_depth); + auto [cols, stubs] = preprocess_columns( + elements, expected_hierarchy, expected_depth, stream, mr.get_temporary_mr()); // generate offsets size_type count = 0; @@ -1628,7 +1860,8 @@ class lists_column_wrapper : public detail::column_wrapper { // add the final offset offsetv.push_back(count); auto offsets = - cudf::test::fixed_width_column_wrapper(offsetv.begin(), offsetv.end()).release(); + cudf::test::fixed_width_column_wrapper(offsetv.begin(), offsetv.end(), stream, mr) + .release(); // concatenate them together, skipping children that are null. std::vector children; @@ -1639,16 +1872,14 @@ class lists_column_wrapper : public detail::column_wrapper { cuda::std::identity{}); auto data = children.empty() ? cudf::empty_like(expected_hierarchy) - : cudf::concatenate(children, - cudf::test::get_default_stream(), - cudf::get_current_device_resource_ref()); + : cudf::concatenate(children, stream, mr.get_output_mr()); // increment depth depth = expected_depth + 1; auto [null_mask, null_count] = [&] { if (v.size() <= 0) return std::make_pair(rmm::device_buffer{}, cudf::size_type{0}); - return cudf::test::detail::make_null_mask(v.begin(), v.end()); + return cudf::test::detail::make_null_mask(v.begin(), v.end(), stream, mr); }(); // construct the list column @@ -1661,9 +1892,13 @@ class lists_column_wrapper : public detail::column_wrapper { * will be "unwrapped" when used in the nesting (list of lists) case. * * @param c Input column to be wrapped + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column * */ - void build_from_non_nested(std::unique_ptr c) + void build_from_non_nested(std::unique_ptr c, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { CUDF_EXPECTS(c->type().id() == type_id::EMPTY || !cudf::is_nested(c->type()), "Unexpected type"); @@ -1674,7 +1909,8 @@ class lists_column_wrapper : public detail::column_wrapper { offsetv.push_back(c->size()); } auto offsets = - cudf::test::fixed_width_column_wrapper(offsetv.begin(), offsetv.end()).release(); + cudf::test::fixed_width_column_wrapper(offsetv.begin(), offsetv.end(), stream, mr) + .release(); // construct the list column. mark this as a root root = true; @@ -1716,11 +1952,15 @@ class lists_column_wrapper : public detail::column_wrapper { * * @param col Input column to be normalized * @param expected_hierarchy Input column which represents the expected hierarchy + * @param stream CUDA stream used for device memory operations + * @param temp_mr Device memory resource used for temporary normalized copies * * @return A new column representing a normalized copy of col */ std::unique_ptr normalize_column(column_view const& col, - column_view const& expected_hierarchy) + column_view const& expected_hierarchy, + rmm::cuda_stream_view stream, + rmm::device_async_resource_ref temp_mr) { // if are at the bottom of the short column, it must be empty if (col.type().id() != type_id::LIST) { @@ -1731,20 +1971,22 @@ class lists_column_wrapper : public detail::column_wrapper { } lists_column_view lcv(col); - return make_lists_column( - col.size(), - std::make_unique(lcv.offsets()), - normalize_column(lists_column_view(col).child(), - lists_column_view(expected_hierarchy).child()), - col.null_count(), - cudf::copy_bitmask( - col, cudf::test::get_default_stream(), cudf::get_current_device_resource_ref())); + return make_lists_column(col.size(), + std::make_unique(lcv.offsets(), stream, temp_mr), + normalize_column(lists_column_view(col).child(), + lists_column_view(expected_hierarchy).child(), + stream, + temp_mr), + col.null_count(), + cudf::copy_bitmask(col, stream, temp_mr)); } std::pair, std::vector>> preprocess_columns( std::initializer_list> const& elements, column_view& expected_hierarchy, - int expected_depth) + int expected_depth, + rmm::cuda_stream_view stream, + rmm::device_async_resource_ref temp_mr) { std::vector> stubs; std::vector cols; @@ -1752,37 +1994,38 @@ class lists_column_wrapper : public detail::column_wrapper { // preprocess the incoming lists. // - unwrap any "root" lists // - handle incomplete hierarchies - std::transform(elements.begin(), - elements.end(), - std::back_inserter(cols), - [&](lists_column_wrapper const& l) -> column_view { - // depth mismatch. attempt to normalize the short column. - // this function will also catch if this is a legitimately broken - // set of input - if (l.depth < expected_depth) { - if (l.root) { - // this exception distinguishes between the following two cases: - // - // { {{{1, 2, 3}}}, {} } - // In this case, row 0 is a List>>, whereas row 1 is - // just a List<> which is an apparent mismatch. However, because row 1 - // is empty we will allow that to semantically mean - // "a List>> that's empty at the top level" - // - // { {{{1, 2, 3}}}, {4, 5, 6} } - // In this case, row 1 is a concrete List with actual values. - // There is no way to rectify the differences so we will treat it as a - // true column mismatch. - CUDF_EXPECTS(l.wrapped->size() == 0, "Mismatch in column types!"); - stubs.push_back(empty_like(expected_hierarchy)); - } else { - stubs.push_back(normalize_column(l.get_view(), expected_hierarchy)); - } - return *(stubs.back()); - } - // the empty hierarchy case - return l.get_view(); - }); + std::transform( + elements.begin(), + elements.end(), + std::back_inserter(cols), + [&](lists_column_wrapper const& l) -> column_view { + // depth mismatch. attempt to normalize the short column. + // this function will also catch if this is a legitimately broken + // set of input + if (l.depth < expected_depth) { + if (l.root) { + // this exception distinguishes between the following two cases: + // + // { {{{1, 2, 3}}}, {} } + // In this case, row 0 is a List>>, whereas row 1 is + // just a List<> which is an apparent mismatch. However, because row 1 + // is empty we will allow that to semantically mean + // "a List>> that's empty at the top level" + // + // { {{{1, 2, 3}}}, {4, 5, 6} } + // In this case, row 1 is a concrete List with actual values. + // There is no way to rectify the differences so we will treat it as a + // true column mismatch. + CUDF_EXPECTS(l.wrapped->size() == 0, "Mismatch in column types!"); + stubs.push_back(empty_like(expected_hierarchy)); + } else { + stubs.push_back(normalize_column(l.get_view(), expected_hierarchy, stream, temp_mr)); + } + return *(stubs.back()); + } + // the empty hierarchy case + return l.get_view(); + }); return {std::move(cols), std::move(stubs)}; } @@ -1796,6 +2039,12 @@ class lists_column_wrapper : public detail::column_wrapper { bool root = false; }; +/** + * @brief True when `T` is convertible to `rmm::cuda_stream_view`. + */ +template +concept convertible_to_cuda_stream_view = std::is_convertible_v; + /** * @brief `column_wrapper` derived class for wrapping columns of structs. */ @@ -1825,13 +2074,24 @@ class structs_column_wrapper : public detail::column_wrapper { * auto struct_col {structs_col.release()}; * @endcode * + * The existing allocations in adopted child columns retain their original memory-resource + * provenance. The supplied output resource controls the struct null mask and any child + * allocations created while sanitizing null struct rows. + * + * To pass an explicit stream/mr with no parent nulls, pass an empty validity: + * `structs_column_wrapper(std::move(children), {}, stream, mr)`. + * * @param child_columns The vector of pre-constructed child columns * @param validity The vector of bools representing the column validity values + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used for new allocations owned by the returned column */ structs_column_wrapper(std::vector>&& child_columns, - std::vector const& validity = {}) + std::vector const& validity = {}, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { - init(std::move(child_columns), validity); + init(std::move(child_columns), validity, stream, mr); } /** @@ -1851,12 +2111,22 @@ class structs_column_wrapper : public detail::column_wrapper { * auto struct_col {structs_col.release()}; * @endcode * + * Child wrappers are deep-copied, so all allocations in the returned children use the supplied + * output resource. The source wrappers retain their original allocations. + * + * To pass an explicit stream/mr with no parent nulls, pass an empty validity: + * `structs_column_wrapper({wrappers}, {}, stream, mr)`. + * * @param child_column_wrappers The list of child column wrappers * @param validity The vector of bools representing the column validity values + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ structs_column_wrapper( std::initializer_list> child_column_wrappers, - std::vector const& validity = {}) + std::vector const& validity = {}, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { std::vector> child_columns; child_columns.reserve(child_column_wrappers.size()); @@ -1864,10 +2134,10 @@ class structs_column_wrapper : public detail::column_wrapper { child_column_wrappers.end(), std::back_inserter(child_columns), [&](auto const& column_wrapper) { - return std::make_unique(column_wrapper.get(), - cudf::test::get_default_stream()); + return std::make_unique( + column_wrapper.get(), stream, mr.get_output_mr()); }); - init(std::move(child_columns), validity); + init(std::move(child_columns), validity, stream, mr); } /** @@ -1889,11 +2159,16 @@ class structs_column_wrapper : public detail::column_wrapper { * * @param child_column_wrappers The list of child column wrappers * @param validity_iter Iterator returning the per-row validity bool + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template structs_column_wrapper( std::initializer_list> child_column_wrappers, - V validity_iter) + V validity_iter, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) + requires(!convertible_to_cuda_stream_view) { std::vector> child_columns; child_columns.reserve(child_column_wrappers.size()); @@ -1901,15 +2176,17 @@ class structs_column_wrapper : public detail::column_wrapper { child_column_wrappers.end(), std::back_inserter(child_columns), [&](auto const& column_wrapper) { - return std::make_unique(column_wrapper.get(), - cudf::test::get_default_stream()); + return std::make_unique( + column_wrapper.get(), stream, mr.get_output_mr()); }); - init(std::move(child_columns), validity_iter); + init(std::move(child_columns), validity_iter, stream, mr); } private: void init(std::vector>&& child_columns, - std::vector const& validity) + std::vector const& validity, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { size_type num_rows = child_columns.empty() ? 0 : child_columns[0]->size(); @@ -1923,19 +2200,22 @@ class structs_column_wrapper : public detail::column_wrapper { auto [null_mask, null_count] = [&] { if (validity.size() <= 0) return std::make_pair(rmm::device_buffer{}, cudf::size_type{0}); - return cudf::test::detail::make_null_mask(validity.begin(), validity.end()); + return cudf::test::detail::make_null_mask(validity.begin(), validity.end(), stream, mr); }(); wrapped = cudf::make_structs_column(num_rows, std::move(child_columns), null_count, std::move(null_mask), - cudf::test::get_default_stream(), - cudf::get_current_device_resource_ref()); + stream, + mr.get_output_mr()); } template - void init(std::vector>&& child_columns, V validity_iterator) + void init(std::vector>&& child_columns, + V validity_iterator, + rmm::cuda_stream_view stream, + cudf::memory_resources mr) { size_type const num_rows = child_columns.empty() ? 0 : child_columns[0]->size(); @@ -1947,7 +2227,7 @@ class structs_column_wrapper : public detail::column_wrapper { std::vector validity(num_rows); std::copy(validity_iterator, validity_iterator + num_rows, validity.begin()); - init(std::move(child_columns), validity); + init(std::move(child_columns), validity, stream, mr); } }; diff --git a/cpp/include/cudf_test/memory_resource_utilities.hpp b/cpp/include/cudf_test/memory_resource_utilities.hpp index 4c63e6741c52..1ad81d7ef1eb 100644 --- a/cpp/include/cudf_test/memory_resource_utilities.hpp +++ b/cpp/include/cudf_test/memory_resource_utilities.hpp @@ -5,6 +5,7 @@ #pragma once +#include #include #include @@ -17,6 +18,7 @@ #include #include +#include #include #include #include @@ -154,6 +156,24 @@ class memory_resource_test_harness { rmm::mr::callback_memory_resource _failing_mr; }; +/** + * @brief Callable that accepts a statistics resource and returns a column wrapper. + */ +template +concept column_wrapper_statistics_resource_factory = + requires(Factory& factory, rmm::mr::statistics_resource_adaptor& mr) { + { std::invoke(factory, mr) } -> std::derived_from; + }; + +/** + * @brief Callable that accepts `cudf::memory_resources` and returns a column wrapper. + */ +template +concept column_wrapper_memory_resources_factory = + requires(Factory& factory, cudf::memory_resources mr) { + { std::invoke(factory, mr) } -> std::derived_from; + }; + /** * @brief Verify that an owning result uses one explicitly supplied output resource. * @@ -169,7 +189,7 @@ class memory_resource_test_harness { * @param output_expectation Expected relationship between live and total output bytes * @param stream Stream to synchronize before inspecting allocation counters */ -template +template void expect_output_uses_resource( Factory&& factory, output_allocation_expectation output_expectation = output_allocation_expectation::EXACT, @@ -207,7 +227,7 @@ void expect_output_uses_resource( * @param expectations Expected output and temporary allocation behavior * @param stream Stream to synchronize before inspecting allocation counters */ -template +template void expect_output_uses_distinct_resources( Factory&& factory, memory_resource_expectations expectations = {}, diff --git a/cpp/include/cudf_test/timestamp_utilities.cuh b/cpp/include/cudf_test/timestamp_utilities.cuh index 7065e5bcce42..c96a8d0dba06 100644 --- a/cpp/include/cudf_test/timestamp_utilities.cuh +++ b/cpp/include/cudf_test/timestamp_utilities.cuh @@ -6,6 +6,7 @@ #pragma once #include +#include #include #include @@ -31,11 +32,16 @@ using time_point_ms = * @param count The number of timestamps to create * @param start The first timestamp as a cuda::std::chrono::time_point * @param stop The last timestamp as a cuda::std::chrono::time_point + * @param stream CUDA stream used for device memory operations + * @param mr Memory resources used to allocate the returned column */ template -inline cudf::test::fixed_width_column_wrapper generate_timestamps(int32_t count, - time_point_ms start, - time_point_ms stop) +inline cudf::test::fixed_width_column_wrapper generate_timestamps( + int32_t count, + time_point_ms start, + time_point_ms stop, + rmm::cuda_stream_view stream = cudf::test::get_default_stream(), + cudf::memory_resources mr = cudf::get_current_device_resource_ref()) { using Rep = typename T::rep; using Period = typename T::period; @@ -56,10 +62,10 @@ inline cudf::test::fixed_width_column_wrapper generate_timestamps(in if (nullable) { auto mask = cudf::detail::make_counting_transform_iterator(0, [](auto i) { return i % 2 == 0; }); - return cudf::test::fixed_width_column_wrapper(iter, iter + count, mask); + return cudf::test::fixed_width_column_wrapper(iter, iter + count, mask, stream, mr); } else { // This needs to be in an else to quash `statement_not_reachable` warnings - return cudf::test::fixed_width_column_wrapper(iter, iter + count); + return cudf::test::fixed_width_column_wrapper(iter, iter + count, stream, mr); } } diff --git a/cpp/tests/utilities/column_utilities.cu b/cpp/tests/utilities/column_utilities.cu index 046ebcbb917b..5d6b2664a651 100644 --- a/cpp/tests/utilities/column_utilities.cu +++ b/cpp/tests/utilities/column_utilities.cu @@ -490,8 +490,8 @@ std::string stringify_column_differences(cudf::device_span difference buffer << depth_str << "differences:" << std::endl; auto source_table = cudf::table_view({lhs, rhs}); - auto diff_column = - fixed_width_column_wrapper(h_differences.begin(), h_differences.end()); + auto diff_column = fixed_width_column_wrapper( + h_differences.begin(), h_differences.end(), stream, mr.get_temporary_mr()); auto diff_table = cudf::gather(source_table, diff_column, cudf::out_of_bounds_policy::DONT_CHECK, @@ -542,6 +542,8 @@ struct column_comparator_impl { auto lhs_tview = table_view{{lhs}}; auto rhs_tview = table_view{{rhs}}; + // TODO: Pass `mr` once two_table_comparator / equality preprocessed_table::create accept + // memory_resources instead of allocating from the current device resource. auto const comparator = cudf::detail::row::equality::two_table_comparator{lhs_tview, rhs_tview, stream}; auto const has_nulls = cudf::has_nulls(lhs_tview) or cudf::has_nulls(rhs_tview); @@ -883,6 +885,9 @@ bool expect_columns_equal(cudf::column_view const& lhs, cuda::stream_ref stream, cudf::memory_resources mr) { + // TODO: equality row preprocessing (two_table_comparator / preprocessed_table::create) still + // allocates from the current device resource; pass `mr` through once that path accepts + // memory_resources so callers need not disable failing current-resource scopes. check_non_empty_nulls(lhs, rhs, stream); auto lhs_indices = generate_all_row_indices(lhs.size(), stream, mr); auto rhs_indices = generate_all_row_indices(rhs.size(), stream, mr); diff --git a/cpp/tests/utilities_tests/column_utilities_tests.cpp b/cpp/tests/utilities_tests/column_utilities_tests.cpp index aa8f988b591f..ae55a56948bc 100644 --- a/cpp/tests/utilities_tests/column_utilities_tests.cpp +++ b/cpp/tests/utilities_tests/column_utilities_tests.cpp @@ -48,21 +48,27 @@ TYPED_TEST_SUITE(ColumnUtilitiesTestFixedPoint, cudf::test::FixedPointTypes); TYPED_TEST(ColumnUtilitiesTest, NonNullableToHost) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + auto sequence = cudf::detail::make_counting_transform_iterator( 0, [](auto i) { return cudf::test::make_type_param_scalar(i); }); auto size = this->size(); std::vector data(sequence, sequence + size); - cudf::test::fixed_width_column_wrapper col(data.begin(), data.end()); + cudf::test::fixed_width_column_wrapper col(data.begin(), data.end(), stream, mr); - auto host_data = cudf::test::to_host(col); + auto host_data = cudf::test::to_host(col, stream, mr); EXPECT_TRUE(std::equal(data.begin(), data.end(), host_data.first.begin())); } TYPED_TEST(ColumnUtilitiesTest, NonNullableToHostWithOffset) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + auto sequence = cudf::detail::make_counting_transform_iterator( 0, [](auto i) { return cudf::test::make_type_param_scalar(i); }); @@ -71,18 +77,22 @@ TYPED_TEST(ColumnUtilitiesTest, NonNullableToHostWithOffset) auto data = std::vector(sequence, sequence + size); auto expected_data = std::vector(sequence + split, sequence + size); - auto col = cudf::test::fixed_width_column_wrapper(data.begin(), data.end()); + auto col = + cudf::test::fixed_width_column_wrapper(data.begin(), data.end(), stream, mr); auto const splits = std::vector{split}; - auto result = cudf::split(col, splits); + auto result = cudf::split(col, splits, stream); - auto host_data = cudf::test::to_host(result.back()); + auto host_data = cudf::test::to_host(result.back(), stream, mr); EXPECT_TRUE(std::equal(expected_data.begin(), expected_data.end(), host_data.first.begin())); } TYPED_TEST(ColumnUtilitiesTest, NullableToHostWithOffset) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + auto sequence = cudf::detail::make_counting_transform_iterator( 0, [](auto i) { return cudf::test::make_type_param_scalar(i); }); @@ -92,12 +102,13 @@ TYPED_TEST(ColumnUtilitiesTest, NullableToHostWithOffset) 0, [&split](auto i) { return i <= 10 and i > split; }); std::vector data(sequence, sequence + size); std::vector expected_data(sequence + split, sequence + size); - cudf::test::fixed_width_column_wrapper col(data.begin(), data.end(), valid); + cudf::test::fixed_width_column_wrapper col( + data.begin(), data.end(), valid, stream, mr); std::vector splits{split}; - std::vector result = cudf::split(col, splits); + std::vector result = cudf::split(col, splits, stream); - auto host_data = cudf::test::to_host(result.back()); + auto host_data = cudf::test::to_host(result.back(), stream, mr); EXPECT_TRUE(std::equal(expected_data.begin(), expected_data.end(), host_data.first.begin())); @@ -108,6 +119,9 @@ TYPED_TEST(ColumnUtilitiesTest, NullableToHostWithOffset) TYPED_TEST(ColumnUtilitiesTest, NullableToHostAllValid) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + auto sequence = cudf::detail::make_counting_transform_iterator( 0, [](auto i) { return cudf::test::make_type_param_scalar(i); }); @@ -116,9 +130,10 @@ TYPED_TEST(ColumnUtilitiesTest, NullableToHostAllValid) auto size = this->size(); std::vector data(sequence, sequence + size); - cudf::test::fixed_width_column_wrapper col(data.begin(), data.end(), all_valid); + cudf::test::fixed_width_column_wrapper col( + data.begin(), data.end(), all_valid, stream, mr); - auto host_data = cudf::test::to_host(col); + auto host_data = cudf::test::to_host(col, stream, mr); EXPECT_TRUE(std::equal(data.begin(), data.end(), host_data.first.begin())); @@ -131,19 +146,28 @@ struct ColumnUtilitiesEquivalenceTest : public cudf::test::BaseFixture {}; TEST_F(ColumnUtilitiesEquivalenceTest, DoubleTest) { - cudf::test::fixed_width_column_wrapper col1{10. / 3, 22. / 7}; - cudf::test::fixed_width_column_wrapper col2{31. / 3 - 21. / 3, 19. / 7 + 3. / 7}; + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + + cudf::test::fixed_width_column_wrapper col1({10. / 3, 22. / 7}, stream, mr); + cudf::test::fixed_width_column_wrapper col2( + {31. / 3 - 21. / 3, 19. / 7 + 3. / 7}, stream, mr); - CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(col1, col2); + CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT( + col1, col2, cudf::test::debug_output_level::FIRST_ERROR, cudf::test::default_ulp, stream, mr); } TEST_F(ColumnUtilitiesEquivalenceTest, NullabilityTest) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + auto all_valid = cudf::test::iterators::no_nulls(); - cudf::test::fixed_width_column_wrapper col1{1, 2, 3}; - cudf::test::fixed_width_column_wrapper col2({1, 2, 3}, all_valid); + cudf::test::fixed_width_column_wrapper col1({1, 2, 3}, stream, mr); + cudf::test::fixed_width_column_wrapper col2({1, 2, 3}, all_valid, stream, mr); - CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(col1, col2); + CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT( + col1, col2, cudf::test::debug_output_level::FIRST_ERROR, cudf::test::default_ulp, stream, mr); } TEST_F(ColumnUtilitiesEquivalenceTest, DistinctMemoryResources) @@ -202,12 +226,17 @@ struct ColumnUtilitiesStringsTest : public cudf::test::BaseFixture {}; TEST_F(ColumnUtilitiesStringsTest, StringsToHost) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + std::vector h_strings{"eee", "bb", nullptr, "", "aa", "bbb", "ééé"}; cudf::test::strings_column_wrapper strings( h_strings.begin(), h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); - auto host_data = cudf::test::to_host(strings); + thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), + stream, + mr); + auto host_data = cudf::test::to_host(strings, stream, mr); auto result_itr = host_data.first.begin(); for (auto itr = h_strings.begin(); itr != h_strings.end(); ++itr, ++result_itr) { if (*itr) { EXPECT_TRUE((*result_itr) == (*itr)); } @@ -216,12 +245,17 @@ TEST_F(ColumnUtilitiesStringsTest, StringsToHost) TEST_F(ColumnUtilitiesStringsTest, StringsToHostAllNulls) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + std::vector h_strings{nullptr, nullptr, nullptr}; cudf::test::strings_column_wrapper strings( h_strings.begin(), h_strings.end(), - thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; })); - auto host_data = cudf::test::to_host(strings); + thrust::make_transform_iterator(h_strings.begin(), [](auto str) { return str != nullptr; }), + stream, + mr); + auto host_data = cudf::test::to_host(strings, stream, mr); auto results = host_data.first; EXPECT_EQ(std::size_t{3}, host_data.first.size()); EXPECT_TRUE(std::all_of(results.begin(), results.end(), [](auto s) { return s.empty(); })); @@ -229,6 +263,9 @@ TEST_F(ColumnUtilitiesStringsTest, StringsToHostAllNulls) TYPED_TEST(ColumnUtilitiesTestFixedPoint, NonNullableToHost) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + using namespace numeric; using decimalXX = TypeParam; using rep = cudf::device_storage_type_t; @@ -239,16 +276,20 @@ TYPED_TEST(ColumnUtilitiesTestFixedPoint, NonNullableToHost) auto fps = cudf::detail::make_counting_transform_iterator(0, to_fp); auto reps = cudf::detail::make_counting_transform_iterator(0, to_rep); - auto const size = 1000; - auto const expected = std::vector(fps, fps + size); - auto const col = cudf::test::fixed_point_column_wrapper(reps, reps + size, scale); - auto const host_data = cudf::test::to_host(col); + auto const size = 1000; + auto const expected = std::vector(fps, fps + size); + auto const col = + cudf::test::fixed_point_column_wrapper(reps, reps + size, scale, stream, mr); + auto const host_data = cudf::test::to_host(col, stream, mr); EXPECT_TRUE(std::equal(expected.begin(), expected.end(), host_data.first.begin())); } TYPED_TEST(ColumnUtilitiesTestFixedPoint, NonNullableToHostWithOffset) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + using namespace numeric; using decimalXX = TypeParam; using rep = cudf::device_storage_type_t; @@ -263,11 +304,12 @@ TYPED_TEST(ColumnUtilitiesTestFixedPoint, NonNullableToHostWithOffset) auto const split = cudf::size_type{2}; auto const expected = std::vector(fps + split, fps + size); - auto const col = cudf::test::fixed_point_column_wrapper(reps, reps + size, scale); - auto const splits = std::vector{split}; - auto result = cudf::split(col, splits); + auto const col = + cudf::test::fixed_point_column_wrapper(reps, reps + size, scale, stream, mr); + auto const splits = std::vector{split}; + auto result = cudf::split(col, splits, stream); - auto host_data = cudf::test::to_host(result.back()); + auto host_data = cudf::test::to_host(result.back(), stream, mr); EXPECT_TRUE(std::equal(expected.begin(), expected.end(), host_data.first.begin())); } @@ -276,65 +318,82 @@ struct ColumnUtilitiesListsTest : public cudf::test::BaseFixture {}; TEST_F(ColumnUtilitiesListsTest, Equivalence) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + // list, nullable vs. non-nullable { auto all_valid = cudf::test::iterators::no_nulls(); - cudf::test::lists_column_wrapper a{{1, 2, 3}, {5, 6}, {8, 9}, {10}, {14, 15}}; - cudf::test::lists_column_wrapper b{{{1, 2, 3}, {5, 6}, {8, 9}, {10}, {14, 15}}, all_valid}; + cudf::test::lists_column_wrapper a( + {{1, 2, 3}, {5, 6}, {8, 9}, {10}, {14, 15}}, stream, mr); + cudf::test::lists_column_wrapper b( + {{1, 2, 3}, {5, 6}, {8, 9}, {10}, {14, 15}}, all_valid, stream, mr); // properties - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT(a, b); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT( + a, b, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); EXPECT_FALSE(cudf::test::detail::expect_column_properties_equal( - a, b, cudf::test::debug_output_level::QUIET)); + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); // values - CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(a, b); - EXPECT_FALSE( - cudf::test::detail::expect_columns_equal(a, b, cudf::test::debug_output_level::QUIET)); + CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT( + a, b, cudf::test::debug_output_level::FIRST_ERROR, cudf::test::default_ulp, stream, mr); + EXPECT_FALSE(cudf::test::detail::expect_columns_equal( + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); } // list>, nullable vs. non-nullable { auto all_valid = cudf::test::iterators::no_nulls(); - cudf::test::lists_column_wrapper a{{{1, 2, 3}, {5, 6}}, {{8, 9}, {10}}, {{14, 15}}}; - cudf::test::lists_column_wrapper b{{{{1, 2, 3}, {5, 6}}, {{8, 9}, {10}}, {{14, 15}}}, - all_valid}; + cudf::test::lists_column_wrapper a( + {{{1, 2, 3}, {5, 6}}, {{8, 9}, {10}}, {{14, 15}}}, stream, mr); + cudf::test::lists_column_wrapper b( + {{{1, 2, 3}, {5, 6}}, {{8, 9}, {10}}, {{14, 15}}}, all_valid, stream, mr); // properties - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT(a, b); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT( + a, b, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); EXPECT_FALSE(cudf::test::detail::expect_column_properties_equal( - a, b, cudf::test::debug_output_level::QUIET)); + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); - CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(a, b); - EXPECT_FALSE( - cudf::test::detail::expect_columns_equal(a, b, cudf::test::debug_output_level::QUIET)); + CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT( + a, b, cudf::test::debug_output_level::FIRST_ERROR, cudf::test::default_ulp, stream, mr); + EXPECT_FALSE(cudf::test::detail::expect_columns_equal( + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); } } TEST_F(ColumnUtilitiesListsTest, DifferingRowCounts) { - cudf::test::fixed_width_column_wrapper a{1, 1, 1, 1}; - cudf::test::fixed_width_column_wrapper b{1, 1, 1, 1, 1}; + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + + cudf::test::fixed_width_column_wrapper a({1, 1, 1, 1}, stream, mr); + cudf::test::fixed_width_column_wrapper b({1, 1, 1, 1, 1}, stream, mr); - EXPECT_FALSE( - cudf::test::detail::expect_columns_equal(a, b, cudf::test::debug_output_level::QUIET)); + EXPECT_FALSE(cudf::test::detail::expect_columns_equal( + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); EXPECT_FALSE(cudf::test::detail::expect_column_properties_equal( - a, b, cudf::test::debug_output_level::QUIET)); - EXPECT_FALSE( - cudf::test::detail::expect_columns_equivalent(a, b, cudf::test::debug_output_level::QUIET)); + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); + EXPECT_FALSE(cudf::test::detail::expect_columns_equivalent( + a, b, cudf::test::debug_output_level::QUIET, cudf::test::default_ulp, stream, mr)); EXPECT_FALSE(cudf::test::detail::expect_column_properties_equivalent( - a, b, cudf::test::debug_output_level::QUIET)); + a, b, cudf::test::debug_output_level::QUIET, stream, mr)); } TEST_F(ColumnUtilitiesListsTest, DifferentPhysicalStructureBeforeConstruction) { + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + // list { std::vector valids = {0, 0, 1, 0, 1, 0, 0}; - cudf::test::fixed_width_column_wrapper c0_offsets{0, 3, 6, 8, 11, 14, 16, 19}; - cudf::test::fixed_width_column_wrapper c0_data{ - 1, 1, 1, 2, 2, 2, 3, 3, 4, 4, 4, 5, 5, 5, 6, 6, 7, 7, 7}; + cudf::test::fixed_width_column_wrapper c0_offsets( + {0, 3, 6, 8, 11, 14, 16, 19}, stream, mr); + cudf::test::fixed_width_column_wrapper c0_data( + {1, 1, 1, 2, 2, 2, 3, 3, 4, 4, 4, 5, 5, 5, 6, 6, 7, 7, 7}, stream, mr); auto [null_mask, null_count] = cudf::test::detail::make_null_mask(valids.begin(), valids.end()); @@ -344,8 +403,8 @@ TEST_F(ColumnUtilitiesListsTest, DifferentPhysicalStructureBeforeConstruction) return cudf::purge_nonempty_nulls(tmp->view()); }(); - cudf::test::fixed_width_column_wrapper c1_offsets{0, 0, 0, 2, 2, 5, 5, 5}; - cudf::test::fixed_width_column_wrapper c1_data{3, 3, 5, 5, 5}; + cudf::test::fixed_width_column_wrapper c1_offsets({0, 0, 0, 2, 2, 5, 5, 5}, stream, mr); + cudf::test::fixed_width_column_wrapper c1_data({3, 3, 5, 5, 5}, stream, mr); auto c1 = [&] { auto tmp = make_lists_column( 7, @@ -357,23 +416,26 @@ TEST_F(ColumnUtilitiesListsTest, DifferentPhysicalStructureBeforeConstruction) }(); // properties - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(*c0, *c1); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL( + *c0, *c1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); // values - CUDF_TEST_EXPECT_COLUMNS_EQUAL(*c0, *c1); + CUDF_TEST_EXPECT_COLUMNS_EQUAL( + *c0, *c1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); } // list>> { std::vector level1_valids = {0, 0, 1, 0, 1, 0, 0}; - cudf::test::fixed_width_column_wrapper c0_l1_offsets{0, 1, 2, 4, 4, 7, 7, 7}; - cudf::test::fixed_width_column_wrapper c0_l2_offsets{0, 1, 2, 5, 6, 7, 10, 14}; - cudf::test::fixed_width_column_wrapper c0_l3_ints{ - 1, 1, -1, -2, -3, 1, 1, -4, -5, -6, -7, -8, -9, -10}; - cudf::test::fixed_width_column_wrapper c0_l3_floats{ - 1, 1, 10, 20, 30, 1, 1, 40, 50, 60, 70, 80, 90, 100}; - cudf::test::structs_column_wrapper c0_l2_data({c0_l3_ints, c0_l3_floats}); + cudf::test::fixed_width_column_wrapper c0_l1_offsets({0, 1, 2, 4, 4, 7, 7, 7}, stream, mr); + cudf::test::fixed_width_column_wrapper c0_l2_offsets( + {0, 1, 2, 5, 6, 7, 10, 14}, stream, mr); + cudf::test::fixed_width_column_wrapper c0_l3_ints( + {1, 1, -1, -2, -3, 1, 1, -4, -5, -6, -7, -8, -9, -10}, stream, mr); + cudf::test::fixed_width_column_wrapper c0_l3_floats( + {1, 1, 10, 20, 30, 1, 1, 40, 50, 60, 70, 80, 90, 100}, stream, mr); + cudf::test::structs_column_wrapper c0_l2_data({c0_l3_ints, c0_l3_floats}, {}, stream, mr); std::vector c0_l2_valids = {1, 1, 1, 0, 0, 1, 1}; auto [null_mask, null_count] = @@ -392,12 +454,13 @@ TEST_F(ColumnUtilitiesListsTest, DifferentPhysicalStructureBeforeConstruction) return cudf::purge_nonempty_nulls(tmp->view()); }(); - cudf::test::fixed_width_column_wrapper c1_l1_offsets{0, 0, 0, 2, 2, 5, 5, 5}; - cudf::test::fixed_width_column_wrapper c1_l2_offsets{0, 3, 3, 3, 6, 10}; - cudf::test::fixed_width_column_wrapper c1_l3_ints{-1, -2, -3, -4, -5, -6, -7, -8, -9, -10}; - cudf::test::fixed_width_column_wrapper c1_l3_floats{ - 10, 20, 30, 40, 50, 60, 70, 80, 90, 100}; - cudf::test::structs_column_wrapper c1_l2_data({c1_l3_ints, c1_l3_floats}); + cudf::test::fixed_width_column_wrapper c1_l1_offsets({0, 0, 0, 2, 2, 5, 5, 5}, stream, mr); + cudf::test::fixed_width_column_wrapper c1_l2_offsets({0, 3, 3, 3, 6, 10}, stream, mr); + cudf::test::fixed_width_column_wrapper c1_l3_ints( + {-1, -2, -3, -4, -5, -6, -7, -8, -9, -10}, stream, mr); + cudf::test::fixed_width_column_wrapper c1_l3_floats( + {10, 20, 30, 40, 50, 60, 70, 80, 90, 100}, stream, mr); + cudf::test::structs_column_wrapper c1_l2_data({c1_l3_ints, c1_l3_floats}, {}, stream, mr); std::vector c1_l2_valids = {1, 0, 0, 1, 1}; std::tie(null_mask, null_count) = @@ -417,10 +480,12 @@ TEST_F(ColumnUtilitiesListsTest, DifferentPhysicalStructureBeforeConstruction) }(); // properties - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(*c0, *c1); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL( + *c0, *c1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); // values - CUDF_TEST_EXPECT_COLUMNS_EQUAL(*c0, *c1); + CUDF_TEST_EXPECT_COLUMNS_EQUAL( + *c0, *c1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); } } @@ -428,62 +493,80 @@ struct ColumnUtilitiesStructsTest : public cudf::test::BaseFixture {}; TEST_F(ColumnUtilitiesStructsTest, Properties) { - cudf::test::strings_column_wrapper s0_scol0{"mno", "jkl", "ghi", "def", "abc"}; - cudf::test::fixed_width_column_wrapper s0_scol1{5, 4, 3, 2, 1}; - cudf::test::strings_column_wrapper s0_sscol0{"5555", "4444", "333", "22", "1"}; - cudf::test::fixed_width_column_wrapper s0_sscol1{50, 40, 30, 20, 10}; - cudf::test::lists_column_wrapper s0_sscol2{{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}; - cudf::test::structs_column_wrapper s0_scol2({s0_sscol0, s0_sscol1, s0_sscol2}); - cudf::test::structs_column_wrapper s_col0({s0_scol0, s0_scol1, s0_scol2}); + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + + cudf::test::strings_column_wrapper s0_scol0({"mno", "jkl", "ghi", "def", "abc"}, stream, mr); + cudf::test::fixed_width_column_wrapper s0_scol1({5, 4, 3, 2, 1}, stream, mr); + cudf::test::strings_column_wrapper s0_sscol0({"5555", "4444", "333", "22", "1"}, stream, mr); + cudf::test::fixed_width_column_wrapper s0_sscol1({50, 40, 30, 20, 10}, stream, mr); + cudf::test::lists_column_wrapper s0_sscol2( + {{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, stream, mr); + cudf::test::structs_column_wrapper s0_scol2({s0_sscol0, s0_sscol1, s0_sscol2}, {}, stream, mr); + cudf::test::structs_column_wrapper s_col0({s0_scol0, s0_scol1, s0_scol2}, {}, stream, mr); auto all_valid = cuda::make_constant_iterator(true); - cudf::test::strings_column_wrapper s1_scol0{"mno", "jkl", "ghi", "def", "abc"}; - cudf::test::fixed_width_column_wrapper s1_scol1{5, 4, 3, 2, 1}; - cudf::test::strings_column_wrapper s1_sscol0{"5555", "4444", "333", "22", "1"}; - cudf::test::fixed_width_column_wrapper s1_sscol1{50, 40, 30, 20, 10}; - cudf::test::lists_column_wrapper s1_sscol2{{{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, - all_valid}; - cudf::test::structs_column_wrapper s1_scol2({s1_sscol0, s1_sscol1, s1_sscol2}); - cudf::test::structs_column_wrapper s_col1({s1_scol0, s1_scol1, s1_scol2}); + cudf::test::strings_column_wrapper s1_scol0({"mno", "jkl", "ghi", "def", "abc"}, stream, mr); + cudf::test::fixed_width_column_wrapper s1_scol1({5, 4, 3, 2, 1}, stream, mr); + cudf::test::strings_column_wrapper s1_sscol0({"5555", "4444", "333", "22", "1"}, stream, mr); + cudf::test::fixed_width_column_wrapper s1_sscol1({50, 40, 30, 20, 10}, stream, mr); + cudf::test::lists_column_wrapper s1_sscol2( + {{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, all_valid, stream, mr); + cudf::test::structs_column_wrapper s1_scol2({s1_sscol0, s1_sscol1, s1_sscol2}, {}, stream, mr); + cudf::test::structs_column_wrapper s_col1({s1_scol0, s1_scol1, s1_scol2}, {}, stream, mr); // equivalent, but not equal - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT(s_col0, s_col1); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUIVALENT( + s_col0, s_col1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); EXPECT_FALSE(cudf::test::detail::expect_column_properties_equal( - s_col0, s_col1, cudf::test::debug_output_level::QUIET)); + s_col0, s_col1, cudf::test::debug_output_level::QUIET, stream, mr)); - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(s_col0, s_col0); - CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL(s_col1, s_col1); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL( + s_col0, s_col0, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); + CUDF_TEST_EXPECT_COLUMN_PROPERTIES_EQUAL( + s_col1, s_col1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); } TEST_F(ColumnUtilitiesStructsTest, Values) { - cudf::test::strings_column_wrapper s0_scol0{"mno", "jkl", "ghi", "def", "abc"}; - cudf::test::fixed_width_column_wrapper s0_scol1{5, 4, 3, 2, 1}; - cudf::test::strings_column_wrapper s0_sscol0{"5555", "4444", "333", "22", "1"}; - cudf::test::fixed_width_column_wrapper s0_sscol1{50, 40, 30, 20, 10}; - cudf::test::lists_column_wrapper s0_sscol2{{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}; - cudf::test::structs_column_wrapper s0_scol2({s0_sscol0, s0_sscol1, s0_sscol2}); - cudf::test::structs_column_wrapper s_col0({s0_scol0, s0_scol1, s0_scol2}); + auto stream = cudf::test::get_default_stream(); + auto mr = this->mr(); + + cudf::test::strings_column_wrapper s0_scol0({"mno", "jkl", "ghi", "def", "abc"}, stream, mr); + cudf::test::fixed_width_column_wrapper s0_scol1({5, 4, 3, 2, 1}, stream, mr); + cudf::test::strings_column_wrapper s0_sscol0({"5555", "4444", "333", "22", "1"}, stream, mr); + cudf::test::fixed_width_column_wrapper s0_sscol1({50, 40, 30, 20, 10}, stream, mr); + cudf::test::lists_column_wrapper s0_sscol2( + {{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, stream, mr); + cudf::test::structs_column_wrapper s0_scol2({s0_sscol0, s0_sscol1, s0_sscol2}, {}, stream, mr); + cudf::test::structs_column_wrapper s_col0({s0_scol0, s0_scol1, s0_scol2}, {}, stream, mr); auto all_valid = cuda::make_constant_iterator(true); - cudf::test::strings_column_wrapper s1_scol0{"mno", "jkl", "ghi", "def", "abc"}; - cudf::test::fixed_width_column_wrapper s1_scol1{5, 4, 3, 2, 1}; - cudf::test::strings_column_wrapper s1_sscol0{"5555", "4444", "333", "22", "1"}; - cudf::test::fixed_width_column_wrapper s1_sscol1{50, 40, 30, 20, 10}; - cudf::test::lists_column_wrapper s1_sscol2{{{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, - all_valid}; - cudf::test::structs_column_wrapper s1_scol2({s1_sscol0, s1_sscol1, s1_sscol2}); - cudf::test::structs_column_wrapper s_col1({s1_scol0, s1_scol1, s1_scol2}); + cudf::test::strings_column_wrapper s1_scol0({"mno", "jkl", "ghi", "def", "abc"}, stream, mr); + cudf::test::fixed_width_column_wrapper s1_scol1({5, 4, 3, 2, 1}, stream, mr); + cudf::test::strings_column_wrapper s1_sscol0({"5555", "4444", "333", "22", "1"}, stream, mr); + cudf::test::fixed_width_column_wrapper s1_sscol1({50, 40, 30, 20, 10}, stream, mr); + cudf::test::lists_column_wrapper s1_sscol2( + {{1, 2}, {3, 4}, {5}, {6, 7, 8}, {12, 12}}, all_valid, stream, mr); + cudf::test::structs_column_wrapper s1_scol2({s1_sscol0, s1_sscol1, s1_sscol2}, {}, stream, mr); + cudf::test::structs_column_wrapper s_col1({s1_scol0, s1_scol1, s1_scol2}, {}, stream, mr); // equivalent, but not equal - CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(s_col0, s_col1); + CUDF_TEST_EXPECT_COLUMNS_EQUIVALENT(s_col0, + s_col1, + cudf::test::debug_output_level::FIRST_ERROR, + cudf::test::default_ulp, + stream, + mr); EXPECT_FALSE(cudf::test::detail::expect_columns_equal( - s_col0, s_col1, cudf::test::debug_output_level::QUIET)); + s_col0, s_col1, cudf::test::debug_output_level::QUIET, stream, mr)); - CUDF_TEST_EXPECT_COLUMNS_EQUAL(s_col0, s_col0); - CUDF_TEST_EXPECT_COLUMNS_EQUAL(s_col1, s_col1); + CUDF_TEST_EXPECT_COLUMNS_EQUAL( + s_col0, s_col0, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); + CUDF_TEST_EXPECT_COLUMNS_EQUAL( + s_col1, s_col1, cudf::test::debug_output_level::FIRST_ERROR, stream, mr); } CUDF_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/utilities_tests/column_wrapper_tests.cpp b/cpp/tests/utilities_tests/column_wrapper_tests.cpp index 34a58ec6184c..dfc5638e2653 100644 --- a/cpp/tests/utilities_tests/column_wrapper_tests.cpp +++ b/cpp/tests/utilities_tests/column_wrapper_tests.cpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -7,15 +7,219 @@ #include #include #include +#include #include #include #include +#include #include +using cudf::test::expect_output_uses_distinct_resources; +using cudf::test::temporary_allocation_expectation; + +namespace { +auto const uses_temporary = cudf::test::memory_resource_expectations{ + cudf::test::output_allocation_expectation::EXACT, temporary_allocation_expectation::SOME}; +} // namespace + +TEST(FixedPointColumnWrapperMemoryResourceTest, DistinctOutputAndTemporaryResources) +{ + auto stream = cudf::test::get_default_stream(); + auto const elements = std::vector{1, 2, 3, 4}; + auto const validity = std::vector{true, false, true, false}; + auto const scale = numeric::scale_type{-2}; + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + elements.begin(), elements.end(), scale, stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + {1, 2, 3, 4}, scale, stream, mr.get_output_mr()); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + elements.begin(), elements.end(), validity.begin(), scale, stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + {1, 2, 3, 4}, {true, false, true, false}, scale, stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + {1, 2, 3, 4}, validity.begin(), scale, stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::fixed_point_column_wrapper( + elements.begin(), elements.end(), {true, false, true, false}, scale, stream, mr); + }); +} + +TEST(StringsColumnWrapperMemoryResourceTest, DistinctOutputAndTemporaryResources) +{ + auto stream = cudf::test::get_default_stream(); + auto const strings = std::vector{"", "alpha", "beta", "gamma"}; + auto const validity = std::vector{true, false, true, false}; + + expect_output_uses_distinct_resources( + [&]([[maybe_unused]] auto mr) { return cudf::test::strings_column_wrapper{}; }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::strings_column_wrapper(strings.begin(), strings.end(), stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::strings_column_wrapper( + {"", "alpha", "beta", "gamma"}, stream, mr.get_output_mr()); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::strings_column_wrapper( + strings.begin(), strings.end(), validity.begin(), stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::strings_column_wrapper( + {"", "alpha", "beta", "gamma"}, validity.begin(), stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::strings_column_wrapper( + {"", "alpha", "beta", "gamma"}, {true, false, true, false}, stream, mr); + }); + + expect_output_uses_distinct_resources([&](auto mr) { + using pair_type = std::pair; + return cudf::test::strings_column_wrapper( + {pair_type{"", true}, pair_type{"alpha", false}, pair_type{"beta", true}}, stream, mr); + }); +} + +TEST(DictionaryColumnWrapperMemoryResourceTest, FixedWidthDistinctOutputAndTemporaryResources) +{ + auto stream = cudf::test::get_default_stream(); + auto const elements = std::vector{3, 1, 3, 2}; + auto const validity = std::vector{true, false, true, true}; + + // Intermediate fixed-width column is allocated on temporary_mr before encode. + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + elements.begin(), elements.end(), stream, mr); + }, + uses_temporary); + + // Single-ref overload: temporaries go to the current resource, not the harness temporary. + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::dictionary_column_wrapper({3, 1, 3, 2}, stream, mr.get_output_mr()); + }); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + elements.begin(), elements.end(), validity.begin(), stream, mr); + }, + uses_temporary); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + {3, 1, 3, 2}, validity.begin(), stream, mr); + }, + uses_temporary); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + {3, 1, 3, 2}, {true, false, true, true}, stream, mr); + }, + uses_temporary); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + elements.begin(), elements.end(), {true, false, true, true}, stream, mr); + }, + uses_temporary); +} + +TEST(DictionaryColumnWrapperMemoryResourceTest, EmptyStringDictionaryPreservesChildTypes) +{ + expect_output_uses_distinct_resources([&]([[maybe_unused]] auto mr) { + auto wrapper = cudf::test::dictionary_column_wrapper(); + auto dictionary = cudf::dictionary_column_view{static_cast(wrapper)}; + + EXPECT_EQ(0, static_cast(wrapper).size()); + EXPECT_EQ(cudf::type_id::STRING, dictionary.keys().type().id()); + EXPECT_EQ(cudf::type_id::INT32, dictionary.indices().type().id()); + return wrapper; + }); +} + +TEST(DictionaryColumnWrapperMemoryResourceTest, StringDistinctOutputAndTemporaryResources) +{ + auto stream = cudf::test::get_default_stream(); + auto const strings = std::vector{"gamma", "alpha", "gamma", "beta"}; + auto const validity = std::vector{true, false, true, true}; + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + strings.begin(), strings.end(), stream, mr); + }, + uses_temporary); + + // Single-ref overload: temporaries go to the current resource, not the harness temporary. + expect_output_uses_distinct_resources([&](auto mr) { + return cudf::test::dictionary_column_wrapper( + {"gamma", "alpha", "gamma", "beta"}, stream, mr.get_output_mr()); + }); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + strings.begin(), strings.end(), validity.begin(), stream, mr); + }, + uses_temporary); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + {"gamma", "alpha", "gamma", "beta"}, validity.begin(), stream, mr); + }, + uses_temporary); + + expect_output_uses_distinct_resources( + [&](auto mr) { + return cudf::test::dictionary_column_wrapper( + {"gamma", "alpha", "gamma", "beta"}, {true, false, true, true}, stream, mr); + }, + uses_temporary); +} + +struct ColumnWrapperTestWithHarness : public cudf::test::BaseFixtureWithHarness { + /** + * @brief Validate that the harness owns the given result. + * + * Assert that the harness output resource holds bytes equal to `col->alloc_size()` and that no + * temporary allocations remain live. `col` is destroyed on return, so `TearDown` can additionally + * confirm that the output bytes were released. + */ + void validate_with_harness(std::unique_ptr col) + { + _harness.expect_resource_usage(col->alloc_size(), {}, this->stream()); + } +}; + template -struct FixedWidthColumnWrapperTest : public cudf::test::BaseFixture, +struct FixedWidthColumnWrapperTest : public ColumnWrapperTestWithHarness, cudf::test::UniformRandomGenerator { FixedWidthColumnWrapperTest() : cudf::test::UniformRandomGenerator{1000, 5000} {} @@ -30,7 +234,7 @@ TYPED_TEST(FixedWidthColumnWrapperTest, EmptyIterator) { auto sequence = cuda::counting_iterator{0}; cudf::test::fixed_width_column_wrapper col( - sequence, sequence); + sequence, sequence, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), 0); EXPECT_EQ(view.head(), nullptr); @@ -38,6 +242,8 @@ TYPED_TEST(FixedWidthColumnWrapperTest, EmptyIterator) EXPECT_FALSE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, EmptyList) { @@ -49,6 +255,8 @@ TYPED_TEST(FixedWidthColumnWrapperTest, EmptyList) EXPECT_FALSE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NonNullableIteratorConstructor) @@ -58,7 +266,7 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NonNullableIteratorConstructor) auto size = this->size(); cudf::test::fixed_width_column_wrapper col( - sequence, sequence + size); + sequence, sequence + size, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), size); EXPECT_NE(nullptr, view.head()); @@ -66,11 +274,14 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NonNullableIteratorConstructor) EXPECT_FALSE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NonNullableListConstructor) { - cudf::test::fixed_width_column_wrapper col({1, 2, 3, 4, 5}); + cudf::test::fixed_width_column_wrapper col( + {1, 2, 3, 4, 5}, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), 5); @@ -79,6 +290,8 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NonNullableListConstructor) EXPECT_FALSE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllValid) @@ -90,7 +303,7 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllValid) auto size = this->size(); cudf::test::fixed_width_column_wrapper col( - sequence, sequence + size, all_valid); + sequence, sequence + size, all_valid, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), size); EXPECT_NE(nullptr, view.head()); @@ -98,13 +311,16 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllValid) EXPECT_TRUE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullableListConstructorAllValid) { auto all_valid = cudf::test::iterators::no_nulls(); - cudf::test::fixed_width_column_wrapper col({1, 2, 3, 4, 5}, all_valid); + cudf::test::fixed_width_column_wrapper col( + {1, 2, 3, 4, 5}, all_valid, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), 5); EXPECT_NE(nullptr, view.head()); @@ -112,6 +328,8 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableListConstructorAllValid) EXPECT_TRUE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllNull) @@ -123,7 +341,7 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllNull) auto size = this->size(); cudf::test::fixed_width_column_wrapper col( - sequence, sequence + size, all_null); + sequence, sequence + size, all_null, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), size); EXPECT_NE(nullptr, view.head()); @@ -132,13 +350,16 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableIteratorConstructorAllNull) EXPECT_TRUE(view.has_nulls()); EXPECT_EQ(view.null_count(), size); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullableListConstructorAllNull) { auto all_null = cudf::test::iterators::all_nulls(); - cudf::test::fixed_width_column_wrapper col({1, 2, 3, 4, 5}, all_null); + cudf::test::fixed_width_column_wrapper col( + {1, 2, 3, 4, 5}, all_null, this->stream(), this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), 5); EXPECT_NE(nullptr, view.head()); @@ -147,13 +368,17 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullableListConstructorAllNull) EXPECT_TRUE(view.has_nulls()); EXPECT_EQ(view.null_count(), 5); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullablePairListConstructorAllNull) { using p = std::pair; cudf::test::fixed_width_column_wrapper col( - {p{1, false}, p{2, false}, p{3, false}, p{4, false}, p{5, false}}); + {p{1, false}, p{2, false}, p{3, false}, p{4, false}, p{5, false}}, + this->stream(), + this->resources()); cudf::column_view view = col; EXPECT_EQ(view.size(), 5); @@ -163,13 +388,16 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullablePairListConstructorAllNull) EXPECT_TRUE(view.has_nulls()); EXPECT_EQ(view.null_count(), 5); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(FixedWidthColumnWrapperTest, NullablePairListConstructorAllNullMatch) { auto odd_valid = cudf::test::iterators::nulls_at_multiples_of(2); - cudf::test::fixed_width_column_wrapper match_col({1, 2, 3, 4, 5}, odd_valid); + cudf::test::fixed_width_column_wrapper match_col( + {1, 2, 3, 4, 5}, odd_valid, this->stream(), this->resources()); cudf::column_view match_view = match_col; using p = std::pair; @@ -177,17 +405,26 @@ TYPED_TEST(FixedWidthColumnWrapperTest, NullablePairListConstructorAllNullMatch) p{2, odd_valid[1]}, p{3, odd_valid[2]}, p{4, odd_valid[3]}, - p{5, odd_valid[4]}}); + p{5, odd_valid[4]}}, + this->stream(), + this->resources()); cudf::column_view view = col; - CUDF_TEST_EXPECT_COLUMNS_EQUAL(view, match_view); + // TODO: Check the harness with a failing current resource once equality row preprocessing uses + // the supplied memory resources. + CUDF_TEST_EXPECT_COLUMNS_EQUAL(view, + match_view, + cudf::test::debug_output_level::FIRST_ERROR, + this->stream(), + this->resources()); } TYPED_TEST(FixedWidthColumnWrapperTest, ReleaseWrapperAllValid) { auto all_valid = cudf::test::iterators::no_nulls(); - cudf::test::fixed_width_column_wrapper col({1, 2, 3, 4, 5}, all_valid); + cudf::test::fixed_width_column_wrapper col( + {1, 2, 3, 4, 5}, all_valid, this->stream(), this->resources()); auto colPtr = col.release(); cudf::column_view view = *colPtr; EXPECT_EQ(view.size(), 5); @@ -196,13 +433,16 @@ TYPED_TEST(FixedWidthColumnWrapperTest, ReleaseWrapperAllValid) EXPECT_TRUE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(std::move(colPtr)); } TYPED_TEST(FixedWidthColumnWrapperTest, ReleaseWrapperAllNull) { auto all_null = cudf::test::iterators::all_nulls(); - cudf::test::fixed_width_column_wrapper col({1, 2, 3, 4, 5}, all_null); + cudf::test::fixed_width_column_wrapper col( + {1, 2, 3, 4, 5}, all_null, this->stream(), this->resources()); auto colPtr = col.release(); cudf::column_view view = *colPtr; EXPECT_EQ(view.size(), 5); @@ -212,11 +452,12 @@ TYPED_TEST(FixedWidthColumnWrapperTest, ReleaseWrapperAllNull) EXPECT_TRUE(view.has_nulls()); EXPECT_EQ(view.null_count(), 5); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(std::move(colPtr)); } template -struct StringsColumnWrapperTest : public cudf::test::BaseFixture, - cudf::test::UniformRandomGenerator { +struct StringsColumnWrapperTest : public ColumnWrapperTestWithHarness { auto data_type() { return cudf::data_type{cudf::type_to_id()}; } }; @@ -232,13 +473,17 @@ TYPED_TEST(StringsColumnWrapperTest, EmptyList) EXPECT_FALSE(view.nullable()); EXPECT_FALSE(view.has_nulls()); EXPECT_EQ(view.offset(), 0); + + this->validate_with_harness(col.release()); } TYPED_TEST(StringsColumnWrapperTest, NullablePairListConstructorAllNull) { using p = std::pair; cudf::test::strings_column_wrapper col( - {p{"a", false}, p{"string", false}, p{"test", false}, p{"for", false}, p{"nulls", false}}); + {p{"a", false}, p{"string", false}, p{"test", false}, p{"for", false}, p{"nulls", false}}, + this->stream(), + this->resources()); cudf::strings_column_view view = cudf::column_view(col); constexpr auto count = 5; @@ -249,14 +494,16 @@ TYPED_TEST(StringsColumnWrapperTest, NullablePairListConstructorAllNull) EXPECT_NE(nullptr, view.offsets().head()); EXPECT_TRUE(view.has_nulls()); EXPECT_EQ(view.null_count(), 5); + + this->validate_with_harness(col.release()); } TYPED_TEST(StringsColumnWrapperTest, NullablePairListConstructorAllNullMatch) { auto odd_valid = cudf::test::iterators::nulls_at_multiples_of(2); - cudf::test::strings_column_wrapper match_col({"a", "string", "", "test", "for", "nulls"}, - odd_valid); + cudf::test::strings_column_wrapper match_col( + {"a", "string", "", "test", "for", "nulls"}, odd_valid, this->stream(), this->resources()); cudf::column_view match_view = match_col; using p = std::pair; @@ -265,8 +512,16 @@ TYPED_TEST(StringsColumnWrapperTest, NullablePairListConstructorAllNullMatch) p{"", odd_valid[2]}, p{"test", odd_valid[3]}, p{"for", odd_valid[4]}, - p{"nulls", odd_valid[5]}}); + p{"nulls", odd_valid[5]}}, + this->stream(), + this->resources()); cudf::column_view view = col; - CUDF_TEST_EXPECT_COLUMNS_EQUAL(view, match_view); + // TODO: Check the harness with a failing current resource once equality row preprocessing uses + // the supplied memory resources. + CUDF_TEST_EXPECT_COLUMNS_EQUAL(view, + match_view, + cudf::test::debug_output_level::FIRST_ERROR, + this->stream(), + this->resources()); } diff --git a/cpp/tests/wrappers/timestamps_test.cu b/cpp/tests/wrappers/timestamps_test.cu index e3af3a50ff54..b91444b6f614 100644 --- a/cpp/tests/wrappers/timestamps_test.cu +++ b/cpp/tests/wrappers/timestamps_test.cu @@ -6,6 +6,7 @@ #include #include #include +#include #include #include #include @@ -64,6 +65,27 @@ struct compare_chrono_elements_to_primitive_representation { }; } // namespace +template +void expect_timestamp_output_uses_resource() +{ + using namespace cuda::std::chrono; + + cudf::test::expect_output_uses_distinct_resources([](auto resources) { + return cudf::test::generate_timestamps( + 100, + cudf::test::time_point_ms{milliseconds{-1000}}, + cudf::test::time_point_ms{milliseconds{1000}}, + cudf::test::get_default_stream(), + resources); + }); +} + +TEST(TimestampGeneratorMemoryResourceTest, DistinctOutputAndTemporaryResources) +{ + expect_timestamp_output_uses_resource(); + expect_timestamp_output_uses_resource(); +} + TYPED_TEST_SUITE(ChronoColumnTest, cudf::test::ChronoTypes); TYPED_TEST(ChronoColumnTest, ChronoDurationsMatchPrimitiveRepresentation)