diff --git a/cpp/include/cudf/detail/utilities/algorithm.cuh b/cpp/include/cudf/detail/utilities/algorithm.cuh index ac9013bbcc1..61a894a474f 100644 --- a/cpp/include/cudf/detail/utilities/algorithm.cuh +++ b/cpp/include/cudf/detail/utilities/algorithm.cuh @@ -6,6 +6,7 @@ #include #include +#include #include #include diff --git a/cpp/include/cudf/detail/utilities/host_vector.hpp b/cpp/include/cudf/detail/utilities/host_vector.hpp index 40ae0a366b5..fbb4166540e 100644 --- a/cpp/include/cudf/detail/utilities/host_vector.hpp +++ b/cpp/include/cudf/detail/utilities/host_vector.hpp @@ -53,10 +53,6 @@ class rmm_host_allocator { }; }; -template -inline constexpr bool contains_property = - (cuda::std::is_same_v || ... || false); - /*! \p rmm_host_allocator is a CUDA-specific host memory allocator * that employs \c `cudf::host_async_resource_ref` for allocation. * @@ -102,11 +98,12 @@ class rmm_host_allocator { #ifdef __CUDACC__ #pragma nv_exec_check_disable #endif - template - rmm_host_allocator(async_host_resource_ref _mr, rmm::cuda_stream_view _stream) + template + rmm_host_allocator(ResourceType _mr, rmm::cuda_stream_view _stream) : mr(_mr), stream(_stream), - _is_device_accessible{contains_property} + _is_device_accessible{ + cuda::mr::synchronous_resource_with} { } diff --git a/cpp/src/io/text/multibyte_split.cu b/cpp/src/io/text/multibyte_split.cu index a11b20db19f..572f59676d7 100644 --- a/cpp/src/io/text/multibyte_split.cu +++ b/cpp/src/io/text/multibyte_split.cu @@ -456,9 +456,10 @@ std::unique_ptr multibyte_split(cudf::io::text::data_chunk_source *thrust::find_if(rmm::exec_policy_nosync(scan_stream), it, it + new_offsets_unclamped, - [row_offsets, byte_range_end] __device__(output_offset i) { - return row_offsets[i] >= byte_range_end; - }); + cuda::proclaim_return_type( + [row_offsets, byte_range_end] __device__(output_offset i) { + return row_offsets[i] >= byte_range_end; + })); // if we had no out-of-bounds offset, we copy all offsets if (end_loc == new_offsets_unclamped) { return end_loc; } // otherwise we copy only up to (including) the first out-of-bounds delimiter diff --git a/cpp/src/join/sort_merge_join.cu b/cpp/src/join/sort_merge_join.cu index 0e8b8f1a86e..a2004a7da5c 100644 --- a/cpp/src/join/sort_merge_join.cu +++ b/cpp/src/join/sort_merge_join.cu @@ -22,6 +22,7 @@ #include #include +#include #include #include #include @@ -124,7 +125,8 @@ merge::matches_per_row(rmm::cuda_stream_view st auto match_counts_it = match_counts.begin(); auto smaller_it = thrust::transform_iterator( sorted_smaller_order_begin, - [] __device__(size_type idx) { return static_cast(idx); }); + cuda::proclaim_return_type( + [] __device__(size_type idx) { return static_cast(idx); })); thrust::upper_bound(rmm::exec_policy_nosync(stream), smaller_it, smaller_it + smaller_numrows, @@ -212,10 +214,12 @@ merge::operator()(rmm::cuda_stream_view stream, }); auto smaller_it = thrust::transform_iterator( sorted_smaller_order_begin, - [] __device__(size_type idx) { return static_cast(idx); }); + cuda::proclaim_return_type( + [] __device__(size_type idx) { return static_cast(idx); })); auto larger_it = thrust::transform_iterator( nonzero_matches.begin(), - [] __device__(size_type idx) { return static_cast(idx); }); + cuda::proclaim_return_type( + [] __device__(size_type idx) { return static_cast(idx); })); thrust::lower_bound(rmm::exec_policy_nosync(stream), smaller_it, smaller_it + smaller_numrows, diff --git a/cpp/src/quantiles/tdigest/tdigest_aggregation.cu b/cpp/src/quantiles/tdigest/tdigest_aggregation.cu index a33cfc65657..903e116494b 100644 --- a/cpp/src/quantiles/tdigest/tdigest_aggregation.cu +++ b/cpp/src/quantiles/tdigest/tdigest_aggregation.cu @@ -738,8 +738,9 @@ cluster_info generate_group_cluster_info(int delta, // CPU. This specifically addresses customer use cases with large inputs and small numbers of // groups, such as just 1. if we're going to be using the CPU, use pinned for a few of the temp // buffers - auto temp_mr = - use_cpu ? cudf::get_pinned_memory_resource() : cudf::get_current_device_resource_ref(); + rmm::device_async_resource_ref temp_mr = + use_cpu ? rmm::device_async_resource_ref{cudf::get_pinned_memory_resource()} + : cudf::get_current_device_resource_ref(); // output from the function cluster_info cinfo; diff --git a/python/cudf/cudf/pandas/scripts/conftest-patch.py b/python/cudf/cudf/pandas/scripts/conftest-patch.py index 0751ce92015..e4ebe09b0f2 100644 --- a/python/cudf/cudf/pandas/scripts/conftest-patch.py +++ b/python/cudf/cudf/pandas/scripts/conftest-patch.py @@ -10454,6 +10454,7 @@ def set_copy_on_write_option(): "tests/indexing/test_iloc.py::TestILocErrors::test_iloc_float_raises[object-Series]", "tests/indexing/test_iloc.py::TestILocErrors::test_iloc_float_raises[uint64-Series]", "tests/indexing/test_iloc.py::TestiLocBaseIndependent::test_setitem_mix_of_nan_and_interval[NaTType-1.0]", + "tests/indexing/test_iloc.py::TestiLocBaseIndependent::test_setitem_mix_of_nan_and_interval[NaTType-a]", "tests/io/excel/test_openpyxl.py::test_engine_kwargs_append_data_only[True-0]", "tests/io/excel/test_writers.py::TestExcelWriterEngineTests::test_ExcelWriter_dispatch[OpenpyxlWriter-.xlsx]", "tests/io/formats/test_format.py::TestDataFrameFormatting::test_auto_detect",