From 50b7f6e64a30e71b0eed96f79ca03629f51e5696 Mon Sep 17 00:00:00 2001 From: Michael Schellenberger Costa Date: Mon, 18 Mar 2024 17:56:51 +0100 Subject: [PATCH 1/5] Replace usages of `thrust::optional` with `cuda::std::optional` --- .../cudf/ast/detail/expression_evaluator.cuh | 8 +-- cpp/include/cudf/ast/detail/operators.hpp | 6 +- .../cudf/column/column_device_view.cuh | 28 ++++----- cpp/include/cudf/detail/copy_if_else.cuh | 6 +- cpp/include/cudf/detail/indexalator.cuh | 12 ++-- cpp/include/cudf/detail/iterator.cuh | 26 ++++---- cpp/include/cudf/json/json.hpp | 2 +- .../strings/detail/convert/fixed_point.cuh | 8 +-- .../cudf/strings/detail/copy_if_else.cuh | 6 +- .../cudf/table/experimental/row_operators.cuh | 6 +- cpp/src/binaryop/binaryop.cpp | 4 +- cpp/src/io/orc/orc.hpp | 7 ++- cpp/src/io/orc/writer_impl.cu | 6 +- .../io/parquet/compact_protocol_reader.cpp | 8 +-- cpp/src/io/parquet/parquet.hpp | 62 +++++++++---------- cpp/src/io/parquet/parquet_gpu.hpp | 14 ++--- cpp/src/io/parquet/predicate_pushdown.cpp | 6 +- cpp/src/io/parquet/reader_impl.cpp | 2 +- cpp/src/io/parquet/reader_impl_chunking.cu | 6 +- cpp/src/io/parquet/reader_impl_helpers.cpp | 6 +- cpp/src/io/parquet/writer_impl.cu | 8 +-- cpp/src/json/json_path.cu | 22 +++---- cpp/src/lists/contains.cu | 2 +- cpp/src/lists/explode.cu | 14 ++--- cpp/src/strings/convert/convert_datetime.cu | 10 +-- .../strings/convert/convert_fixed_point.cu | 1 + cpp/src/strings/regex/regex.cuh | 4 +- cpp/src/strings/regex/regex.inl | 6 +- cpp/src/strings/replace/multi_re.cu | 2 +- cpp/src/transform/row_bit_count.cu | 18 +++--- cpp/tests/io/parquet_common.cpp | 2 +- cpp/tests/io/parquet_common.hpp | 2 +- cpp/tests/iterator/indexalator_test.cu | 11 ++-- cpp/tests/iterator/offsetalator_test.cu | 4 +- cpp/tests/iterator/optional_iterator_test.cuh | 25 ++++---- .../optional_iterator_test_numeric.cu | 10 +-- 36 files changed, 187 insertions(+), 183 deletions(-) diff --git a/cpp/include/cudf/ast/detail/expression_evaluator.cuh b/cpp/include/cudf/ast/detail/expression_evaluator.cuh index 105d87ff96f..683f45fec58 100644 --- a/cpp/include/cudf/ast/detail/expression_evaluator.cuh +++ b/cpp/include/cudf/ast/detail/expression_evaluator.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2023, NVIDIA CORPORATION. + * Copyright (c) 2021-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -29,7 +29,7 @@ #include -#include +#include namespace cudf { @@ -278,7 +278,7 @@ struct expression_evaluator { detail::device_data_reference const& input_reference, IntermediateDataType* thread_intermediate_storage, cudf::size_type left_row_index, - thrust::optional right_row_index = {}) const + cuda::std::optional right_row_index = {}) const { // TODO: Everywhere in the code assumes that the table reference is either // left or right. Should we error-check somewhere to prevent @@ -329,7 +329,7 @@ struct expression_evaluator { detail::device_data_reference const& device_data_reference, IntermediateDataType* thread_intermediate_storage, cudf::size_type left_row_index, - thrust::optional right_row_index = {}) const + cuda::std::optional right_row_index = {}) const { CUDF_UNREACHABLE("Unsupported type in resolve_input."); } diff --git a/cpp/include/cudf/ast/detail/operators.hpp b/cpp/include/cudf/ast/detail/operators.hpp index c483d459833..a757adfcdbd 100644 --- a/cpp/include/cudf/ast/detail/operators.hpp +++ b/cpp/include/cudf/ast/detail/operators.hpp @@ -21,8 +21,8 @@ #include #include +#include #include -#include #include #include @@ -35,14 +35,14 @@ namespace ast { namespace detail { -// Type trait for wrapping nullable types in a thrust::optional. Non-nullable +// Type trait for wrapping nullable types in a cuda::std::optional. Non-nullable // types are returned as is. template struct possibly_null_value; template struct possibly_null_value { - using type = thrust::optional; + using type = cuda::std::optional; }; template diff --git a/cpp/include/cudf/column/column_device_view.cuh b/cpp/include/cudf/column/column_device_view.cuh index 787e9c2c479..755958659af 100644 --- a/cpp/include/cudf/column/column_device_view.cuh +++ b/cpp/include/cudf/column/column_device_view.cuh @@ -32,9 +32,9 @@ #include +#include #include #include -#include #include #include @@ -614,7 +614,7 @@ class alignas(16) column_device_view : public detail::column_device_view_base { /** * @brief Return an optional iterator to the first element of the column. * - * Dereferencing the returned iterator returns a `thrust::optional`. + * Dereferencing the returned iterator returns a `cuda::std::optional`. * * The element of this iterator contextually converts to bool. The conversion returns true * if the object contains a value and false if it does not contain a value. @@ -739,7 +739,7 @@ class alignas(16) column_device_view : public detail::column_device_view_base { /** * @brief Return an optional iterator to the element following the last element of the column. * - * The returned iterator represents a `thrust::optional` element. + * The returned iterator represents a `cuda::std::optional` element. * * This function does not participate in overload resolution if * `column_device_view::has_element_accessor()` is false. @@ -1272,21 +1272,21 @@ struct value_accessor { * @brief optional accessor of a column * * - * The optional_accessor always returns a `thrust::optional` of `column[i]`. The validity + * The optional_accessor always returns a `cuda::std::optional` of `column[i]`. The validity * of the optional is determined by the `Nullate` parameter which may be one of the following: * * - `nullate::YES` means that the column supports nulls and the optional returned * might be valid or invalid. * * - `nullate::NO` means the caller attests that the column has no null values, - * no checks will occur and `thrust::optional{column[i]}` will be + * no checks will occur and `cuda::std::optional{column[i]}` will be * return for each `i`. * * - `nullate::DYNAMIC` defers the assumption of nullability to runtime and the caller * specifies if the column has nulls at runtime. - * For `DYNAMIC{true}` the return value will be `thrust::optional{column[i]}` if - * element `i` is not null and `thrust::optional{}` if element `i` is null. - * For `DYNAMIC{false}` the return value will always be `thrust::optional{column[i]}`. + * For `DYNAMIC{true}` the return value will be `cuda::std::optional{column[i]}` if + * element `i` is not null and `cuda::std::optional{}` if element `i` is null. + * For `DYNAMIC{false}` the return value will always be `cuda::std::optional{column[i]}`. * * @throws cudf::logic_error if column datatype and template T type mismatch. * @throws cudf::logic_error if the column is not nullable and `with_nulls` evaluates to true @@ -1312,19 +1312,19 @@ struct optional_accessor { } /** - * @brief Returns a `thrust::optional` of `column[i]`. + * @brief Returns a `cuda::std::optional` of `column[i]`. * * @param i The index of the element to return - * @return A `thrust::optional` that contains the value of `column[i]` is not null. If that + * @return A `cuda::std::optional` that contains the value of `column[i]` is not null. If that * element is null, the resulting optional will not contain a value. */ - __device__ inline thrust::optional operator()(cudf::size_type i) const + __device__ inline cuda::std::optional operator()(cudf::size_type i) const { if (has_nulls) { - return (col.is_valid_nocheck(i)) ? thrust::optional{col.element(i)} - : thrust::optional{thrust::nullopt}; + return (col.is_valid_nocheck(i)) ? cuda::std::optional{col.element(i)} + : cuda::std::optional{cuda::std::nullopt}; } - return thrust::optional{col.element(i)}; + return cuda::std::optional{col.element(i)}; } Nullate has_nulls{}; ///< Indicates if the `col` should be checked for nulls. diff --git a/cpp/include/cudf/detail/copy_if_else.cuh b/cpp/include/cudf/detail/copy_if_else.cuh index 8418e279ce7..d260a4591b7 100644 --- a/cpp/include/cudf/detail/copy_if_else.cuh +++ b/cpp/include/cudf/detail/copy_if_else.cuh @@ -25,8 +25,8 @@ #include #include +#include #include -#include namespace cudf { namespace detail { @@ -70,7 +70,7 @@ __launch_bounds__(block_size) CUDF_KERNEL while (warp_cur <= warp_end) { auto const index = static_cast(tidx); auto const opt_value = - (index < end) ? (filter(index) ? lhs[index] : rhs[index]) : thrust::nullopt; + (index < end) ? (filter(index) ? lhs[index] : rhs[index]) : cuda::std::nullopt; if (opt_value) { out.element(index) = static_cast(*opt_value); } // update validity @@ -156,7 +156,7 @@ std::unique_ptr copy_if_else(bool nullable, rmm::cuda_stream_view stream, rmm::device_async_resource_ref mr) { - // This is the type of the thrust::optional element in the passed iterators + // This is the type of the cuda::std::optional element in the passed iterators using Element = typename thrust::iterator_traits::value_type::value_type; size_type size = std::distance(lhs_begin, lhs_end); diff --git a/cpp/include/cudf/detail/indexalator.cuh b/cpp/include/cudf/detail/indexalator.cuh index b5d57da6cd5..ec7b1c3e6b6 100644 --- a/cpp/include/cudf/detail/indexalator.cuh +++ b/cpp/include/cudf/detail/indexalator.cuh @@ -22,9 +22,9 @@ #include #include +#include #include #include -#include #include namespace cudf { @@ -376,10 +376,10 @@ struct indexalator_factory { iter = make_input_iterator(col); } - __device__ thrust::optional operator()(size_type i) const + __device__ cuda::std::optional operator()(size_type i) const { - return has_nulls && !bit_is_set(null_mask, i + offset) ? thrust::nullopt - : thrust::make_optional(iter[i]); + return has_nulls && !bit_is_set(null_mask, i + offset) ? cuda::std::nullopt + : cuda::std::make_optional(iter[i]); } }; @@ -400,9 +400,9 @@ struct indexalator_factory { iter = indexalator_factory::make_input_iterator(input); } - __device__ thrust::optional operator()(size_type) const + __device__ cuda::std::optional operator()(size_type) const { - return is_null ? thrust::nullopt : thrust::make_optional(*iter); + return is_null ? cuda::std::nullopt : cuda::std::make_optional(*iter); } }; diff --git a/cpp/include/cudf/detail/iterator.cuh b/cpp/include/cudf/detail/iterator.cuh index 9e6227ec19b..4349e1b70fd 100644 --- a/cpp/include/cudf/detail/iterator.cuh +++ b/cpp/include/cudf/detail/iterator.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2019-2023, NVIDIA CORPORATION. + * Copyright (c) 2019-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -37,10 +37,10 @@ #include #include +#include #include #include #include -#include #include #include @@ -186,7 +186,7 @@ auto make_null_replacement_iterator(column_device_view const& column, /** * @brief Constructs an optional iterator over a column's values and its validity. * - * Dereferencing the returned iterator returns a `thrust::optional`. + * Dereferencing the returned iterator returns a `cuda::std::optional`. * * The element of this iterator contextually converts to bool. The conversion returns true * if the object contains a value and false if it does not contain a value. @@ -237,7 +237,7 @@ auto make_null_replacement_iterator(column_device_view const& column, * @param column The column to iterate * @param has_nulls Indicates whether `column` is checked for nulls. * @return Iterator that returns valid column elements and the validity of the - * element in a `thrust::optional` + * element in a `cuda::std::optional` */ template auto make_optional_iterator(column_device_view const& column, Nullate has_nulls) @@ -393,7 +393,7 @@ auto inline make_scalar_iterator(scalar const& scalar_value) /** * @brief Optional accessor for a scalar * - * The `scalar_optional_accessor` always returns a `thrust::optional` of the scalar. + * The `scalar_optional_accessor` always returns a `cuda::std::optional` of the scalar. * The validity of the optional is determined by the `Nullate` parameter which may * be one of the following: * @@ -401,14 +401,14 @@ auto inline make_scalar_iterator(scalar const& scalar_value) * will contain a value only if the scalar is valid. * * - `nullate::NO` means the caller attests that the scalar will always be valid, - * no checks will occur and `thrust::optional{column[i]}` will return a value + * no checks will occur and `cuda::std::optional{column[i]}` will return a value * for each `i`. * * - `nullate::DYNAMIC` defers the assumption of nullability to runtime and the caller * specifies if the scalar may be valid or invalid. - * For `DYNAMIC{true}` the return value will be a `thrust::optional{scalar}` when the - * scalar is valid and a `thrust::optional{}` when the scalar is invalid. - * For `DYNAMIC{false}` the return value will always be a `thrust::optional{scalar}`. + * For `DYNAMIC{true}` the return value will be a `cuda::std::optional{scalar}` when the + * scalar is valid and a `cuda::std::optional{}` when the scalar is invalid. + * For `DYNAMIC{false}` the return value will always be a `cuda::std::optional{scalar}`. * * @throws `cudf::logic_error` if scalar datatype and Element type mismatch. * @@ -418,7 +418,7 @@ auto inline make_scalar_iterator(scalar const& scalar_value) template struct scalar_optional_accessor : public scalar_value_accessor { using super_t = scalar_value_accessor; - using value_type = thrust::optional; + using value_type = cuda::std::optional; scalar_optional_accessor(scalar const& scalar_value, Nullate with_nulls) : scalar_value_accessor(scalar_value), has_nulls{with_nulls} @@ -427,7 +427,7 @@ struct scalar_optional_accessor : public scalar_value_accessor { __device__ inline value_type const operator()(size_type) const { - if (has_nulls && !super_t::dscalar.is_valid()) { return value_type{thrust::nullopt}; } + if (has_nulls && !super_t::dscalar.is_valid()) { return value_type{cuda::std::nullopt}; } if constexpr (cudf::is_fixed_point()) { using namespace numeric; @@ -519,7 +519,7 @@ struct scalar_representation_pair_accessor : public scalar_value_accessor`. + * Dereferencing the returned iterator returns a `cuda::std::optional`. * * The element of this iterator contextually converts to bool. The conversion returns true * if the object contains a value and false if it does not contain a value. @@ -575,7 +575,7 @@ struct scalar_representation_pair_accessor : public scalar_value_accessor auto inline make_optional_iterator(scalar const& scalar_value, Nullate has_nulls) diff --git a/cpp/include/cudf/json/json.hpp b/cpp/include/cudf/json/json.hpp index 385e8e54bdc..b7349080b7c 100644 --- a/cpp/include/cudf/json/json.hpp +++ b/cpp/include/cudf/json/json.hpp @@ -21,7 +21,7 @@ #include #include -#include +#include namespace cudf { diff --git a/cpp/include/cudf/strings/detail/convert/fixed_point.cuh b/cpp/include/cudf/strings/detail/convert/fixed_point.cuh index 5f51da967d3..8440805960e 100644 --- a/cpp/include/cudf/strings/detail/convert/fixed_point.cuh +++ b/cpp/include/cudf/strings/detail/convert/fixed_point.cuh @@ -17,8 +17,8 @@ #include +#include #include -#include #include namespace cudf { @@ -88,7 +88,7 @@ __device__ inline thrust::pair parse_integer( * @return Integer value of the exponent */ template -__device__ thrust::optional parse_exponent(char const* iter, char const* iter_end) +__device__ cuda::std::optional parse_exponent(char const* iter, char const* iter_end) { constexpr uint32_t exponent_max = static_cast(std::numeric_limits::max()); @@ -105,12 +105,12 @@ __device__ thrust::optional parse_exponent(char const* iter, char const while (iter < iter_end) { auto const ch = *iter++; if (ch < '0' || ch > '9') { - if (check_only) { return thrust::nullopt; } + if (check_only) { return cuda::std::nullopt; } break; } uint32_t exp_check = static_cast(exp_ten * 10) + static_cast(ch - '0'); - if (check_only && (exp_check > exponent_max)) { return thrust::nullopt; } // check overflow + if (check_only && (exp_check > exponent_max)) { return cuda::std::nullopt; } // check overflow exp_ten = static_cast(exp_check); } diff --git a/cpp/include/cudf/strings/detail/copy_if_else.cuh b/cpp/include/cudf/strings/detail/copy_if_else.cuh index 4db7651330b..213a41ca596 100644 --- a/cpp/include/cudf/strings/detail/copy_if_else.cuh +++ b/cpp/include/cudf/strings/detail/copy_if_else.cuh @@ -25,8 +25,8 @@ #include #include +#include #include -#include #include namespace cudf { @@ -41,9 +41,9 @@ namespace detail { * ``` * * @tparam StringIterLeft A random access iterator whose value_type is - * `thrust::optional` where the `optional` has a value iff the element is valid. + * `cuda::std::optional` where the `optional` has a value iff the element is valid. * @tparam StringIterRight A random access iterator whose value_type is - * `thrust::optional` where the `optional` has a value iff the element is valid. + * `cuda::std::optional` where the `optional` has a value iff the element is valid. * @tparam Filter Functor that takes an index and returns a boolean. * * @param lhs_begin Start of first set of data. Used when `filter_fn` returns true. diff --git a/cpp/include/cudf/table/experimental/row_operators.cuh b/cpp/include/cudf/table/experimental/row_operators.cuh index c181ac7d402..91739b2011a 100644 --- a/cpp/include/cudf/table/experimental/row_operators.cuh +++ b/cpp/include/cudf/table/experimental/row_operators.cuh @@ -211,7 +211,7 @@ struct sorting_physical_element_comparator { } }; -using optional_dremel_view = thrust::optional; +using optional_dremel_view = cuda::std::optional; // The has_nested_columns template parameter of the device_row_comparator is // necessary to help the compiler optimize our code. Without it, the list and @@ -223,12 +223,12 @@ using optional_dremel_view = thrust::optional; // std::optional> in the // preprocessed_table/device_row_comparator (which is always valid when // has_nested_columns and is otherwise invalid) that is then unpacked to a -// thrust::optional at the element_comparator level (which +// cuda::std::optional at the element_comparator level (which // is always valid for a list column and otherwise invalid). We cannot use an // additional template parameter for the element_comparator on a per-column // basis because we cannot conditionally define dremel_device_view member // variables without jumping through extra hoops with inheritance, so the -// thrust::optional member must be an optional rather than +// cuda::std::optional member must be an optional rather than // a raw dremel_device_view. /** * @brief Computes the lexicographic comparison between 2 rows. diff --git a/cpp/src/binaryop/binaryop.cpp b/cpp/src/binaryop/binaryop.cpp index 3ac8547baad..25b0f68aaa8 100644 --- a/cpp/src/binaryop/binaryop.cpp +++ b/cpp/src/binaryop/binaryop.cpp @@ -41,7 +41,7 @@ #include #include -#include +#include #include @@ -173,7 +173,7 @@ template void fixed_point_binary_operation_validation(binary_operator op, Lhs lhs, Rhs rhs, - thrust::optional output_type = {}) + cuda::std::optional output_type = {}) { CUDF_EXPECTS((is_fixed_point(lhs) or is_fixed_point(rhs)), "One of the inputs must have fixed_point data_type."); diff --git a/cpp/src/io/orc/orc.hpp b/cpp/src/io/orc/orc.hpp index e1403acd455..790532c9d54 100644 --- a/cpp/src/io/orc/orc.hpp +++ b/cpp/src/io/orc/orc.hpp @@ -24,7 +24,7 @@ #include #include -#include +#include #include #include @@ -692,11 +692,12 @@ class metadata { * @brief `column_device_view` and additional, ORC specific, information on the column. */ struct orc_column_device_view : public column_device_view { - __device__ orc_column_device_view(column_device_view col, thrust::optional parent_idx) + __device__ orc_column_device_view(column_device_view col, + cuda::std::optional parent_idx) : column_device_view{col}, parent_index{parent_idx} { } - thrust::optional parent_index; + cuda::std::optional parent_index; bitmask_type const* pushdown_mask = nullptr; }; diff --git a/cpp/src/io/orc/writer_impl.cu b/cpp/src/io/orc/writer_impl.cu index f3b8cfbc836..04eee68e757 100644 --- a/cpp/src/io/orc/writer_impl.cu +++ b/cpp/src/io/orc/writer_impl.cu @@ -42,6 +42,7 @@ #include #include #include +#include #include #include #include @@ -50,7 +51,6 @@ #include #include #include -#include #include #include #include @@ -1831,7 +1831,7 @@ orc_table_view make_orc_table_view(table_view const& table, type_kinds, stream, rmm::mr::get_current_device_resource()); rmm::device_uvector d_orc_columns(orc_columns.size(), stream); - using stack_value_type = thrust::pair>; + using stack_value_type = thrust::pair>; rmm::device_uvector stack_storage(orc_columns.size(), stream); // pre-order append ORC device columns @@ -1847,7 +1847,7 @@ orc_table_view make_orc_table_view(table_view const& table, thrust::make_reverse_iterator(d_table.end()), thrust::make_reverse_iterator(d_table.begin()), [&stack](column_device_view const& c) { - stack.push({&c, thrust::nullopt}); + stack.push({&c, cuda::std::nullopt}); }); uint32_t idx = 0; diff --git a/cpp/src/io/parquet/compact_protocol_reader.cpp b/cpp/src/io/parquet/compact_protocol_reader.cpp index 192833507b0..7067a719d0f 100644 --- a/cpp/src/io/parquet/compact_protocol_reader.cpp +++ b/cpp/src/io/parquet/compact_protocol_reader.cpp @@ -304,10 +304,10 @@ class parquet_field_struct : public parquet_field { template class parquet_field_union_struct : public parquet_field { E& enum_val; - thrust::optional& val; // union structs are always wrapped in std::optional + cuda::std::optional& val; // union structs are always wrapped in std::optional public: - parquet_field_union_struct(int f, E& ev, thrust::optional& v) + parquet_field_union_struct(int f, E& ev, cuda::std::optional& v) : parquet_field(f), enum_val(ev), val(v) { } @@ -431,10 +431,10 @@ class parquet_field_struct_blob : public parquet_field { */ template class parquet_field_optional : public parquet_field { - thrust::optional& val; + cuda::std::optional& val; public: - parquet_field_optional(int f, thrust::optional& v) : parquet_field(f), val(v) {} + parquet_field_optional(int f, cuda::std::optional& v) : parquet_field(f), val(v) {} inline void operator()(CompactProtocolReader* cpr, int field_type) { diff --git a/cpp/src/io/parquet/parquet.hpp b/cpp/src/io/parquet/parquet.hpp index 8ee4c175e09..5d10472b0ae 100644 --- a/cpp/src/io/parquet/parquet.hpp +++ b/cpp/src/io/parquet/parquet.hpp @@ -20,7 +20,7 @@ #include -#include +#include #include #include @@ -94,10 +94,10 @@ struct LogicalType { BSON }; Type type; - thrust::optional decimal_type; - thrust::optional time_type; - thrust::optional timestamp_type; - thrust::optional int_type; + cuda::std::optional decimal_type; + cuda::std::optional time_type; + cuda::std::optional timestamp_type; + cuda::std::optional int_type; LogicalType(Type tp = UNDEFINED) : type(tp) {} LogicalType(DecimalType&& dt) : type(DECIMAL), decimal_type(dt) {} @@ -178,21 +178,21 @@ struct SchemaElement { // 5: nested fields int32_t num_children = 0; // 6: DEPRECATED: record the original type before conversion to parquet type - thrust::optional converted_type; + cuda::std::optional converted_type; // 7: DEPRECATED: record the scale for DECIMAL converted type int32_t decimal_scale = 0; // 8: DEPRECATED: record the precision for DECIMAL converted type int32_t decimal_precision = 0; // 9: save field_id from original schema - thrust::optional field_id; + cuda::std::optional field_id; // 10: replaces converted type - thrust::optional logical_type; + cuda::std::optional logical_type; // extra cudf specific fields bool output_as_byte_array = false; // cudf type determined from arrow:schema - thrust::optional arrow_type; + cuda::std::optional arrow_type; // The following fields are filled in later during schema initialization int max_definition_level = 0; @@ -259,21 +259,21 @@ struct SchemaElement { */ struct Statistics { // deprecated max value in signed comparison order - thrust::optional> max; + cuda::std::optional> max; // deprecated min value in signed comparison order - thrust::optional> min; + cuda::std::optional> min; // count of null values in the column - thrust::optional null_count; + cuda::std::optional null_count; // count of distinct values occurring - thrust::optional distinct_count; + cuda::std::optional distinct_count; // max value for column determined by ColumnOrder - thrust::optional> max_value; + cuda::std::optional> max_value; // min value for column determined by ColumnOrder - thrust::optional> min_value; + cuda::std::optional> min_value; // If true, max_value is the actual maximum value for a column - thrust::optional is_max_value_exact; + cuda::std::optional is_max_value_exact; // If true, min_value is the actual minimum value for a column - thrust::optional is_min_value_exact; + cuda::std::optional is_min_value_exact; }; /** @@ -282,7 +282,7 @@ struct Statistics { struct SizeStatistics { // Number of variable-width bytes stored for the page/chunk. Should not be set for anything // but the BYTE_ARRAY physical type. - thrust::optional unencoded_byte_array_data_bytes; + cuda::std::optional unencoded_byte_array_data_bytes; /** * When present, there is expected to be one element corresponding to each * repetition (i.e. size=max repetition_level+1) where each element @@ -291,14 +291,14 @@ struct SizeStatistics { * * This value should not be written if max_repetition_level is 0. */ - thrust::optional> repetition_level_histogram; + cuda::std::optional> repetition_level_histogram; /** * Same as repetition_level_histogram except for definition levels. * * This value should not be written if max_definition_level is 0 or 1. */ - thrust::optional> definition_level_histogram; + cuda::std::optional> definition_level_histogram; }; /** @@ -319,7 +319,7 @@ struct OffsetIndex { std::vector page_locations; // per-page size info. see description of the same field in SizeStatistics. only present for // columns with a BYTE_ARRAY physical type. - thrust::optional> unencoded_byte_array_data_bytes; + cuda::std::optional> unencoded_byte_array_data_bytes; }; /** @@ -331,10 +331,10 @@ struct ColumnIndex { std::vector> max_values; // upper bound for values in each page BoundaryOrder boundary_order = BoundaryOrder::UNORDERED; // Indicates if min and max values are ordered - thrust::optional> null_counts; // Optional count of null values per page + cuda::std::optional> null_counts; // Optional count of null values per page // Repetition/definition level histograms for the column chunk - thrust::optional> repetition_level_histogram; - thrust::optional> definition_level_histogram; + cuda::std::optional> repetition_level_histogram; + cuda::std::optional> definition_level_histogram; }; /** @@ -384,11 +384,11 @@ struct ColumnChunkMetaData { Statistics statistics; // Set of all encodings used for pages in this column chunk. This information can be used to // determine if all data pages are dictionary encoded for example. - thrust::optional> encoding_stats; + cuda::std::optional> encoding_stats; // Optional statistics to help estimate total memory when converted to in-memory representations. // The histograms contained in these statistics can also be useful in some cases for more // fine-grained nullability/list length filter pushdown. - thrust::optional size_statistics; + cuda::std::optional size_statistics; }; /** @@ -430,13 +430,13 @@ struct RowGroup { int64_t num_rows = 0; // If set, specifies a sort ordering of the rows in this RowGroup. // The sorting columns can be a subset of all the columns. - thrust::optional> sorting_columns; + cuda::std::optional> sorting_columns; // Byte offset from beginning of file to first page (data or dictionary) in this row group - thrust::optional file_offset; + cuda::std::optional file_offset; // Total byte size of all compressed (and potentially encrypted) column data in this row group - thrust::optional total_compressed_size; + cuda::std::optional total_compressed_size; // Row group ordinal in the file - thrust::optional ordinal; + cuda::std::optional ordinal; }; /** @@ -461,7 +461,7 @@ struct FileMetaData { std::vector row_groups; std::vector key_value_metadata; std::string created_by = ""; - thrust::optional> column_orders; + cuda::std::optional> column_orders; }; /** diff --git a/cpp/src/io/parquet/parquet_gpu.hpp b/cpp/src/io/parquet/parquet_gpu.hpp index efc1f5ebab1..8f52f073833 100644 --- a/cpp/src/io/parquet/parquet_gpu.hpp +++ b/cpp/src/io/parquet/parquet_gpu.hpp @@ -394,7 +394,7 @@ struct ColumnChunkDesc { uint8_t def_level_bits_, uint8_t rep_level_bits_, Compression codec_, - thrust::optional logical_type_, + cuda::std::optional logical_type_, int32_t ts_clock_rate_, int32_t src_col_index_, int32_t src_col_schema_, @@ -438,12 +438,12 @@ struct ColumnChunkDesc { int32_t num_data_pages{}; // number of data pages int32_t num_dict_pages{}; // number of dictionary pages PageInfo const* dict_page{}; - string_index_pair* str_dict_index{}; // index for string dictionary - bitmask_type** valid_map_base{}; // base pointers of valid bit map for this column - void** column_data_base{}; // base pointers of column data - void** column_string_base{}; // base pointers of column string data - Compression codec{}; // compressed codec enum - thrust::optional logical_type{}; // logical type + string_index_pair* str_dict_index{}; // index for string dictionary + bitmask_type** valid_map_base{}; // base pointers of valid bit map for this column + void** column_data_base{}; // base pointers of column data + void** column_string_base{}; // base pointers of column string data + Compression codec{}; // compressed codec enum + cuda::std::optional logical_type{}; // logical type int32_t ts_clock_rate{}; // output timestamp clock frequency (0=default, 1000=ms, 1000000000=ns) int32_t src_col_index{}; // my input column index diff --git a/cpp/src/io/parquet/predicate_pushdown.cpp b/cpp/src/io/parquet/predicate_pushdown.cpp index 481c1e9fcdd..5ca090b05b3 100644 --- a/cpp/src/io/parquet/predicate_pushdown.cpp +++ b/cpp/src/io/parquet/predicate_pushdown.cpp @@ -154,7 +154,7 @@ struct stats_caster { } void set_index(size_type index, - thrust::optional> const& binary_value, + cuda::std::optional> const& binary_value, Type const type) { if (binary_value.has_value()) { @@ -236,8 +236,8 @@ struct stats_caster { max.set_index(stats_idx, max_value, colchunk.meta_data.type); } else { // Marking it null, if column present in row group - min.set_index(stats_idx, thrust::nullopt, {}); - max.set_index(stats_idx, thrust::nullopt, {}); + min.set_index(stats_idx, cuda::std::nullopt, {}); + max.set_index(stats_idx, cuda::std::nullopt, {}); } stats_idx++; } diff --git a/cpp/src/io/parquet/reader_impl.cpp b/cpp/src/io/parquet/reader_impl.cpp index 68ec61ead0a..2648a1f41ab 100644 --- a/cpp/src/io/parquet/reader_impl.cpp +++ b/cpp/src/io/parquet/reader_impl.cpp @@ -39,7 +39,7 @@ namespace { // be treated as a string. Currently the only logical type that has special handling is DECIMAL. // Other valid types in the future would be UUID (still treated as string) and FLOAT16 (which // for now would also be treated as a string). -inline bool is_treat_fixed_length_as_string(thrust::optional const& logical_type) +inline bool is_treat_fixed_length_as_string(cuda::std::optional const& logical_type) { if (!logical_type.has_value()) { return true; } return logical_type->type != LogicalType::DECIMAL; diff --git a/cpp/src/io/parquet/reader_impl_chunking.cu b/cpp/src/io/parquet/reader_impl_chunking.cu index 794750ab6d2..54ba898b058 100644 --- a/cpp/src/io/parquet/reader_impl_chunking.cu +++ b/cpp/src/io/parquet/reader_impl_chunking.cu @@ -370,11 +370,11 @@ int64_t find_next_split(int64_t cur_pos, * * @return A tuple of Parquet clock rate and Parquet decimal type. */ -[[nodiscard]] std::tuple> conversion_info( +[[nodiscard]] std::tuple> conversion_info( type_id column_type_id, type_id timestamp_type_id, Type physical, - thrust::optional logical_type) + cuda::std::optional logical_type) { int32_t const clock_rate = is_chrono(data_type{column_type_id}) ? to_clockrate(timestamp_type_id) : 0; @@ -385,7 +385,7 @@ int64_t find_next_split(int64_t cur_pos, // if decimal but not outputting as float or decimal, then convert to no logical type if (column_type_id != type_id::FLOAT64 and not cudf::is_fixed_point(data_type{column_type_id})) { - return std::make_tuple(clock_rate, thrust::nullopt); + return std::make_tuple(clock_rate, cuda::std::nullopt); } } diff --git a/cpp/src/io/parquet/reader_impl_helpers.cpp b/cpp/src/io/parquet/reader_impl_helpers.cpp index 581c44d024b..00f75e4e828 100644 --- a/cpp/src/io/parquet/reader_impl_helpers.cpp +++ b/cpp/src/io/parquet/reader_impl_helpers.cpp @@ -38,7 +38,7 @@ namespace flatbuf = cudf::io::parquet::flatbuf; namespace { -thrust::optional converted_to_logical_type(SchemaElement const& schema) +cuda::std::optional converted_to_logical_type(SchemaElement const& schema) { if (schema.converted_type.has_value()) { switch (schema.converted_type.value()) { @@ -66,7 +66,7 @@ thrust::optional converted_to_logical_type(SchemaElement const& sch default: return LogicalType{LogicalType::UNDEFINED}; } } - return thrust::nullopt; + return cuda::std::nullopt; } } // namespace @@ -246,7 +246,7 @@ void metadata::sanitize_schema() struct_elem.repetition_type = REQUIRED; struct_elem.num_children = schema_elem.num_children; struct_elem.type = UNDEFINED_TYPE; - struct_elem.converted_type = thrust::nullopt; + struct_elem.converted_type = cuda::std::nullopt; // swap children struct_elem.children_idx = std::move(schema_elem.children_idx); diff --git a/cpp/src/io/parquet/writer_impl.cu b/cpp/src/io/parquet/writer_impl.cu index 2df71b77301..e59dae1cc99 100644 --- a/cpp/src/io/parquet/writer_impl.cu +++ b/cpp/src/io/parquet/writer_impl.cu @@ -184,7 +184,7 @@ struct aggregate_writer_metadata { std::vector> column_indexes; }; std::vector files; - thrust::optional> column_orders = thrust::nullopt; + cuda::std::optional> column_orders = cuda::std::nullopt; }; namespace { @@ -470,7 +470,7 @@ struct leaf_schema_fn { std::enable_if_t, void> operator()() { col_schema.type = (timestamp_is_int96) ? Type::INT96 : Type::INT64; - col_schema.converted_type = thrust::nullopt; + col_schema.converted_type = cuda::std::nullopt; col_schema.stats_dtype = statistics_dtype::dtype_timestamp64; if (timestamp_is_int96) { col_schema.ts_scale = -1000; // negative value indicates division by absolute value @@ -748,7 +748,7 @@ std::vector construct_parquet_schema_tree( col_schema.type = Type::BYTE_ARRAY; } - col_schema.converted_type = thrust::nullopt; + col_schema.converted_type = cuda::std::nullopt; col_schema.stats_dtype = statistics_dtype::dtype_byte_array; col_schema.repetition_type = col_nullable ? OPTIONAL : REQUIRED; col_schema.name = (schema[parent_idx].name == "list") ? "element" : col_meta.get_name(); @@ -2808,7 +2808,7 @@ std::unique_ptr> writer::merge_row_group_metadata( // See https://github.com/rapidsai/cudf/pull/14264#issuecomment-1778311615 for (auto& se : md.schema) { if (se.logical_type.has_value() && se.logical_type.value().type == LogicalType::UNKNOWN) { - se.logical_type = thrust::nullopt; + se.logical_type = cuda::std::nullopt; } } diff --git a/cpp/src/json/json_path.cu b/cpp/src/json/json_path.cu index d1a1097de35..1bf4bf3b153 100644 --- a/cpp/src/json/json_path.cu +++ b/cpp/src/json/json_path.cu @@ -39,7 +39,7 @@ #include #include -#include +#include #include #include #include @@ -207,7 +207,7 @@ class parser { struct json_output { size_t output_max_len; char* output; - thrust::optional output_len; + cuda::std::optional output_len; __device__ void add_output(char const* str, size_t len) { @@ -656,7 +656,7 @@ class path_state : private parser { * @param stream Cuda stream to perform any gpu actions on * @returns A pair containing the command buffer, and maximum stack depth required. */ -std::pair>, int> build_command_buffer( +std::pair>, int> build_command_buffer( cudf::string_scalar const& json_path, rmm::cuda_stream_view stream) { std::string h_json_path = json_path.to_string(stream); @@ -690,8 +690,8 @@ std::pair>, int> build_comma } while (op.type != path_operator_type::END); auto const is_empty = h_operators.size() == 1 && h_operators[0].type == path_operator_type::END; - return is_empty ? std::pair(thrust::nullopt, 0) - : std::pair(thrust::make_optional(cudf::detail::make_device_uvector_sync( + return is_empty ? std::pair(cuda::std::nullopt, 0) + : std::pair(cuda::std::make_optional(cudf::detail::make_device_uvector_sync( h_operators, stream, rmm::mr::get_current_device_resource())), max_stack_depth); } @@ -920,9 +920,9 @@ __launch_bounds__(block_size) CUDF_KERNEL path_operator const* const commands, size_type* d_sizes, cudf::detail::input_offsetalator output_offsets, - thrust::optional out_buf, - thrust::optional out_validity, - thrust::optional out_valid_count, + cuda::std::optional out_buf, + cuda::std::optional out_validity, + cuda::std::optional out_valid_count, get_json_object_options options) { auto tid = cudf::detail::grid_1d::global_thread_id(); @@ -1012,9 +1012,9 @@ std::unique_ptr get_json_object(cudf::strings_column_view const& c std::get<0>(preprocess).value().data(), sizes.data(), d_offsets, - thrust::nullopt, - thrust::nullopt, - thrust::nullopt, + cuda::std::nullopt, + cuda::std::nullopt, + cuda::std::nullopt, options); // convert sizes to offsets diff --git a/cpp/src/lists/contains.cu b/cpp/src/lists/contains.cu index f03d394d6d7..0861b6ed63c 100644 --- a/cpp/src/lists/contains.cu +++ b/cpp/src/lists/contains.cu @@ -34,12 +34,12 @@ #include #include +#include #include #include #include #include #include -#include #include #include #include diff --git a/cpp/src/lists/explode.cu b/cpp/src/lists/explode.cu index 46c4fc78a6f..74a0d842aad 100644 --- a/cpp/src/lists/explode.cu +++ b/cpp/src/lists/explode.cu @@ -29,6 +29,7 @@ #include #include +#include #include #include #include @@ -36,7 +37,6 @@ #include #include #include -#include #include #include @@ -57,8 +57,8 @@ std::unique_ptr build_table( size_type const explode_column_idx, column_view const& sliced_child, cudf::device_span gather_map, - thrust::optional> explode_col_gather_map, - thrust::optional> position_array, + cuda::std::optional> explode_col_gather_map, + cuda::std::optional> position_array, rmm::cuda_stream_view stream, rmm::device_async_resource_ref mr) { @@ -143,8 +143,8 @@ std::unique_ptr
explode(table_view const& input_table, explode_column_idx, sliced_child, gather_map, - thrust::nullopt, - thrust::nullopt, + cuda::std::nullopt, + cuda::std::nullopt, stream, mr); } @@ -193,7 +193,7 @@ std::unique_ptr
explode_position(table_view const& input_table, explode_column_idx, sliced_child, gather_map, - thrust::nullopt, + cuda::std::nullopt, std::move(pos), stream, mr); @@ -292,7 +292,7 @@ std::unique_ptr
explode_outer(table_view const& input_table, sliced_child, gather_map, explode_col_gather_map, - include_position ? std::move(pos) : thrust::optional>{}, + include_position ? std::move(pos) : cuda::std::optional>{}, stream, mr); } diff --git a/cpp/src/strings/convert/convert_datetime.cu b/cpp/src/strings/convert/convert_datetime.cu index 64a2107e17a..99c40f00b00 100644 --- a/cpp/src/strings/convert/convert_datetime.cu +++ b/cpp/src/strings/convert/convert_datetime.cu @@ -36,11 +36,11 @@ #include #include +#include #include #include #include #include -#include #include #include @@ -519,7 +519,7 @@ struct check_datetime_format { * The checking here is a little more strict than the actual * parser used for conversion. */ - __device__ thrust::optional check_string(string_view const& d_string) + __device__ cuda::std::optional check_string(string_view const& d_string) { timestamp_components dateparts = {1970, 1, 1, 0}; // init to epoch time @@ -529,7 +529,7 @@ struct check_datetime_format { // eliminate static character values first if (item.item_type == format_char_type::literal) { // check static character matches - if (*ptr != item.value) return thrust::nullopt; + if (*ptr != item.value) return cuda::std::nullopt; ptr += item.length; length -= item.length; continue; @@ -645,7 +645,7 @@ struct check_datetime_format { case 'Z': result = true; // skip default: break; } - if (!result) return thrust::nullopt; + if (!result) return cuda::std::nullopt; ptr += bytes_read; length -= bytes_read; } @@ -821,7 +821,7 @@ struct datetime_formatter_fn { // We only dissect the timestamp into components if needed // by a specifier. And then we only do it once and reuse it. // This can improve performance when not using uncommon specifiers. - thrust::optional days; + cuda::std::optional days; auto days_from_timestamp = [tstamp]() { auto const count = tstamp.time_since_epoch().count(); diff --git a/cpp/src/strings/convert/convert_fixed_point.cu b/cpp/src/strings/convert/convert_fixed_point.cu index 73089ad407e..ac63cffb48f 100644 --- a/cpp/src/strings/convert/convert_fixed_point.cu +++ b/cpp/src/strings/convert/convert_fixed_point.cu @@ -35,6 +35,7 @@ #include #include +#include #include #include #include diff --git a/cpp/src/strings/regex/regex.cuh b/cpp/src/strings/regex/regex.cuh index e6134296e45..2df404048f7 100644 --- a/cpp/src/strings/regex/regex.cuh +++ b/cpp/src/strings/regex/regex.cuh @@ -23,8 +23,8 @@ #include +#include #include -#include #include #include @@ -36,7 +36,7 @@ namespace detail { struct relist; using match_pair = thrust::pair; -using match_result = thrust::optional; +using match_result = cuda::std::optional; constexpr int32_t MAX_SHARED_MEM = 2048; ///< Memory size for storing prog instruction data constexpr std::size_t MAX_WORKING_MEM = 0x01'FFFF'FFFF; ///< Memory size for state data diff --git a/cpp/src/strings/regex/regex.inl b/cpp/src/strings/regex/regex.inl index 23e1944cda4..3b899e4edc1 100644 --- a/cpp/src/strings/regex/regex.inl +++ b/cpp/src/strings/regex/regex.inl @@ -260,12 +260,12 @@ __device__ __forceinline__ match_result reprog_device::regexec(string_view const switch (jnk.starttype) { case BOL: if (pos == 0) break; - if (jnk.startchar != '^') { return thrust::nullopt; } + if (jnk.startchar != '^') { return cuda::std::nullopt; } --itr; startchar = static_cast('\n'); case CHAR: { auto const find_itr = find_char(startchar, dstr, itr); - if (find_itr.byte_offset() >= dstr.size_bytes()) { return thrust::nullopt; } + if (find_itr.byte_offset() >= dstr.size_bytes()) { return cuda::std::nullopt; } itr = find_itr + (jnk.starttype == BOL); pos = itr.position(); break; @@ -396,7 +396,7 @@ __device__ __forceinline__ match_result reprog_device::regexec(string_view const checkstart = jnk.list1->get_size() == 0; } while (!last_character && (!checkstart || !match)); - return match ? match_result({begin, end}) : thrust::nullopt; + return match ? match_result({begin, end}) : cuda::std::nullopt; } __device__ __forceinline__ match_result reprog_device::find(int32_t const thread_idx, diff --git a/cpp/src/strings/replace/multi_re.cu b/cpp/src/strings/replace/multi_re.cu index 31234ea42ec..0ad3ab2305c 100644 --- a/cpp/src/strings/replace/multi_re.cu +++ b/cpp/src/strings/replace/multi_re.cu @@ -92,7 +92,7 @@ struct replace_multi_regex_fn { } reprog_device prog = progs[ptn_idx]; - auto const result = !prog.is_empty() ? prog.find(idx, d_str, itr) : thrust::nullopt; + auto const result = !prog.is_empty() ? prog.find(idx, d_str, itr) : cuda::std::nullopt; d_ranges[ptn_idx] = result ? found_range{result->first, result->second} : found_range{nchars, nchars}; } diff --git a/cpp/src/transform/row_bit_count.cu b/cpp/src/transform/row_bit_count.cu index bfac7ab586e..a5b59d428bf 100644 --- a/cpp/src/transform/row_bit_count.cu +++ b/cpp/src/transform/row_bit_count.cu @@ -34,8 +34,8 @@ #include #include +#include #include -#include #include namespace cudf { @@ -158,9 +158,9 @@ void flatten_hierarchy(ColIter begin, std::vector& info, hierarchy_info& h_info, rmm::cuda_stream_view stream, - size_type cur_depth = 0, - size_type cur_branch_depth = 0, - thrust::optional parent_index = {}); + size_type cur_depth = 0, + size_type cur_branch_depth = 0, + cuda::std::optional parent_index = {}); /** * @brief Type-dispatched functor called by flatten_hierarchy. @@ -176,7 +176,7 @@ struct flatten_functor { rmm::cuda_stream_view, size_type cur_depth, size_type cur_branch_depth, - thrust::optional) + cuda::std::optional) { out.push_back(col); info.push_back({cur_depth, cur_branch_depth, cur_branch_depth}); @@ -193,7 +193,7 @@ struct flatten_functor { rmm::cuda_stream_view, size_type cur_depth, size_type cur_branch_depth, - thrust::optional) + cuda::std::optional) { out.push_back(col); info.push_back({cur_depth, cur_branch_depth, cur_branch_depth}); @@ -209,7 +209,7 @@ struct flatten_functor { rmm::cuda_stream_view stream, size_type cur_depth, size_type cur_branch_depth, - thrust::optional parent_index) + cuda::std::optional parent_index) { // track branch depth as we reach this list and after we pass it auto const branch_depth_start = cur_branch_depth; @@ -242,7 +242,7 @@ struct flatten_functor { rmm::cuda_stream_view stream, size_type cur_depth, size_type cur_branch_depth, - thrust::optional) + cuda::std::optional) { out.push_back(col); info.push_back({cur_depth, cur_branch_depth, cur_branch_depth}); @@ -283,7 +283,7 @@ void flatten_hierarchy(ColIter begin, rmm::cuda_stream_view stream, size_type cur_depth, size_type cur_branch_depth, - thrust::optional parent_index) + cuda::std::optional parent_index) { std::for_each(begin, end, [&](column_view const& col) { cudf::type_dispatcher(col.type(), diff --git a/cpp/tests/io/parquet_common.cpp b/cpp/tests/io/parquet_common.cpp index c1211869bcc..3dd5ad145ea 100644 --- a/cpp/tests/io/parquet_common.cpp +++ b/cpp/tests/io/parquet_common.cpp @@ -744,7 +744,7 @@ int32_t compare(T& v1, T& v2) int32_t compare_binary(std::vector const& v1, std::vector const& v2, cudf::io::parquet::detail::Type ptype, - thrust::optional const& ctype) + cuda::std::optional const& ctype) { auto ctype_val = ctype.value_or(cudf::io::parquet::detail::UNKNOWN); switch (ptype) { diff --git a/cpp/tests/io/parquet_common.hpp b/cpp/tests/io/parquet_common.hpp index 59ee85444f2..bc6145d77da 100644 --- a/cpp/tests/io/parquet_common.hpp +++ b/cpp/tests/io/parquet_common.hpp @@ -172,7 +172,7 @@ std::pair create_parquet_typed_with_stats(std::string int32_t compare_binary(std::vector const& v1, std::vector const& v2, cudf::io::parquet::detail::Type ptype, - thrust::optional const& ctype); + cuda::std::optional const& ctype); void expect_compression_stats_empty(std::shared_ptr stats); diff --git a/cpp/tests/iterator/indexalator_test.cu b/cpp/tests/iterator/indexalator_test.cu index 0c10853ec02..dac2356dcb0 100644 --- a/cpp/tests/iterator/indexalator_test.cu +++ b/cpp/tests/iterator/indexalator_test.cu @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2023, NVIDIA CORPORATION. + * Copyright (c) 2021-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -20,10 +20,10 @@ #include +#include #include #include #include -#include #include #include #include @@ -84,15 +84,16 @@ TYPED_TEST(IndexalatorTest, optional_iterator) auto d_col = cudf::test::fixed_width_column_wrapper( host_values.begin(), host_values.end(), validity.begin()); - auto expected_values = thrust::host_vector>(host_values.size()); + auto expected_values = + thrust::host_vector>(host_values.size()); std::transform(host_values.begin(), host_values.end(), validity.begin(), expected_values.begin(), [](T v, bool b) { - return (b) ? thrust::make_optional(static_cast(v)) - : thrust::nullopt; + return (b) ? cuda::std::make_optional(static_cast(v)) + : cuda::std::nullopt; }); auto it_dev = cudf::detail::indexalator_factory::make_input_optional_iterator(d_col); diff --git a/cpp/tests/iterator/offsetalator_test.cu b/cpp/tests/iterator/offsetalator_test.cu index e569e58f42a..af12ebbe08e 100644 --- a/cpp/tests/iterator/offsetalator_test.cu +++ b/cpp/tests/iterator/offsetalator_test.cu @@ -1,5 +1,5 @@ /* - * Copyright (c) 2023, NVIDIA CORPORATION. + * Copyright (c) 2023-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -20,10 +20,10 @@ #include +#include #include #include #include -#include #include #include #include diff --git a/cpp/tests/iterator/optional_iterator_test.cuh b/cpp/tests/iterator/optional_iterator_test.cuh index 6a264cee9a8..04f5410a44f 100644 --- a/cpp/tests/iterator/optional_iterator_test.cuh +++ b/cpp/tests/iterator/optional_iterator_test.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2020-2022, NVIDIA CORPORATION. + * Copyright (c) 2020-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -16,8 +16,8 @@ #include +#include #include -#include template void nonull_optional_iterator(IteratorTest& testFixture) @@ -32,9 +32,9 @@ void nonull_optional_iterator(IteratorTest& testFixture) auto d_col = cudf::column_device_view::create(w_col); // calculate the expected value by CPU. - thrust::host_vector> replaced_array(host_values.size()); + thrust::host_vector> replaced_array(host_values.size()); std::transform(host_values.begin(), host_values.end(), replaced_array.begin(), [](auto s) { - return thrust::optional{s}; + return cuda::std::optional{s}; }); // GPU test @@ -61,19 +61,20 @@ void null_optional_iterator(IteratorTest& testFixture) auto d_col = cudf::column_device_view::create(w_col); // calculate the expected value by CPU. - thrust::host_vector> optional_values(host_values.size()); - std::transform(host_values.begin(), - host_values.end(), - host_bools.begin(), - optional_values.begin(), - [](auto s, bool b) { return b ? thrust::optional{s} : thrust::optional{}; }); + thrust::host_vector> optional_values(host_values.size()); + std::transform( + host_values.begin(), + host_values.end(), + host_bools.begin(), + optional_values.begin(), + [](auto s, bool b) { return b ? cuda::std::optional{s} : cuda::std::optional{}; }); - thrust::host_vector> value_all_valid(host_values.size()); + thrust::host_vector> value_all_valid(host_values.size()); std::transform(host_values.begin(), host_values.end(), host_bools.begin(), value_all_valid.begin(), - [](auto s, bool b) { return thrust::optional{s}; }); + [](auto s, bool b) { return cuda::std::optional{s}; }); // GPU test for correct null mapping testFixture.iterator_test_thrust( diff --git a/cpp/tests/iterator/optional_iterator_test_numeric.cu b/cpp/tests/iterator/optional_iterator_test_numeric.cu index 98befb0a3ee..257c0979017 100644 --- a/cpp/tests/iterator/optional_iterator_test_numeric.cu +++ b/cpp/tests/iterator/optional_iterator_test_numeric.cu @@ -18,9 +18,9 @@ #include +#include #include #include -#include #include #include @@ -49,21 +49,21 @@ TYPED_TEST(NumericOptionalIteratorTest, null_optional_iterator) { null_optional_ // Transformers and Operators for optional_iterator test template struct transformer_optional_meanvar { - using ResultType = thrust::optional>; + using ResultType = cuda::std::optional>; - CUDF_HOST_DEVICE inline ResultType operator()(thrust::optional const& optional) + CUDF_HOST_DEVICE inline ResultType operator()(cuda::std::optional const& optional) { if (optional.has_value()) { auto v = *optional; return cudf::meanvar{v, static_cast(v * v), 1}; } - return thrust::nullopt; + return cuda::std::nullopt; } }; template struct optional_to_meanvar { - CUDF_HOST_DEVICE inline T operator()(thrust::optional const& v) { return v.value_or(T{0}); } + CUDF_HOST_DEVICE inline T operator()(cuda::std::optional const& v) { return v.value_or(T{0}); } }; // TODO: enable this test also at __CUDACC_DEBUG__ From a09e8a08fee97b10a5b4c6d51f32c01a7da53134 Mon Sep 17 00:00:00 2001 From: Michael Schellenberger Costa Date: Wed, 24 Jul 2024 20:30:31 +0200 Subject: [PATCH 2/5] Apply suggestions from code review Co-authored-by: David Wendt <45795991+davidwendt@users.noreply.github.com> --- cpp/include/cudf/json/json.hpp | 1 - cpp/src/lists/contains.cu | 1 - cpp/src/strings/convert/convert_fixed_point.cu | 1 - 3 files changed, 3 deletions(-) diff --git a/cpp/include/cudf/json/json.hpp b/cpp/include/cudf/json/json.hpp index b7349080b7c..c6127fcdc39 100644 --- a/cpp/include/cudf/json/json.hpp +++ b/cpp/include/cudf/json/json.hpp @@ -21,7 +21,6 @@ #include #include -#include namespace cudf { diff --git a/cpp/src/lists/contains.cu b/cpp/src/lists/contains.cu index 0861b6ed63c..4f6bf5235a7 100644 --- a/cpp/src/lists/contains.cu +++ b/cpp/src/lists/contains.cu @@ -34,7 +34,6 @@ #include #include -#include #include #include #include diff --git a/cpp/src/strings/convert/convert_fixed_point.cu b/cpp/src/strings/convert/convert_fixed_point.cu index ac63cffb48f..73089ad407e 100644 --- a/cpp/src/strings/convert/convert_fixed_point.cu +++ b/cpp/src/strings/convert/convert_fixed_point.cu @@ -35,7 +35,6 @@ #include #include -#include #include #include #include From 8460e193557afae428894aae0be93d92efa93c90 Mon Sep 17 00:00:00 2001 From: Michael Schellenberger Costa Date: Thu, 25 Jul 2024 16:31:46 +0200 Subject: [PATCH 3/5] Fix formatting --- cpp/include/cudf/json/json.hpp | 1 - 1 file changed, 1 deletion(-) diff --git a/cpp/include/cudf/json/json.hpp b/cpp/include/cudf/json/json.hpp index c6127fcdc39..7c90a0d1fa5 100644 --- a/cpp/include/cudf/json/json.hpp +++ b/cpp/include/cudf/json/json.hpp @@ -21,7 +21,6 @@ #include #include - namespace cudf { /** From 7d17f60c0c657fc1bd6c3d62f020b4ced4ab3f5e Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Mon, 19 Aug 2024 13:43:57 -0500 Subject: [PATCH 4/5] Remove include from offsetalator_test.cu --- cpp/tests/iterator/offsetalator_test.cu | 1 - 1 file changed, 1 deletion(-) diff --git a/cpp/tests/iterator/offsetalator_test.cu b/cpp/tests/iterator/offsetalator_test.cu index af12ebbe08e..b206ff947bb 100644 --- a/cpp/tests/iterator/offsetalator_test.cu +++ b/cpp/tests/iterator/offsetalator_test.cu @@ -20,7 +20,6 @@ #include -#include #include #include #include From 0418c913a5f53ac97268bd5560e8eeb31f679c7d Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Mon, 19 Aug 2024 13:46:19 -0500 Subject: [PATCH 5/5] Revert changes in AST code paths --- cpp/include/cudf/ast/detail/expression_evaluator.cuh | 8 ++++---- cpp/include/cudf/ast/detail/operators.hpp | 6 +++--- 2 files changed, 7 insertions(+), 7 deletions(-) diff --git a/cpp/include/cudf/ast/detail/expression_evaluator.cuh b/cpp/include/cudf/ast/detail/expression_evaluator.cuh index 683f45fec58..105d87ff96f 100644 --- a/cpp/include/cudf/ast/detail/expression_evaluator.cuh +++ b/cpp/include/cudf/ast/detail/expression_evaluator.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2024, NVIDIA CORPORATION. + * Copyright (c) 2021-2023, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -29,7 +29,7 @@ #include -#include +#include namespace cudf { @@ -278,7 +278,7 @@ struct expression_evaluator { detail::device_data_reference const& input_reference, IntermediateDataType* thread_intermediate_storage, cudf::size_type left_row_index, - cuda::std::optional right_row_index = {}) const + thrust::optional right_row_index = {}) const { // TODO: Everywhere in the code assumes that the table reference is either // left or right. Should we error-check somewhere to prevent @@ -329,7 +329,7 @@ struct expression_evaluator { detail::device_data_reference const& device_data_reference, IntermediateDataType* thread_intermediate_storage, cudf::size_type left_row_index, - cuda::std::optional right_row_index = {}) const + thrust::optional right_row_index = {}) const { CUDF_UNREACHABLE("Unsupported type in resolve_input."); } diff --git a/cpp/include/cudf/ast/detail/operators.hpp b/cpp/include/cudf/ast/detail/operators.hpp index 7a5df431584..46507700e21 100644 --- a/cpp/include/cudf/ast/detail/operators.hpp +++ b/cpp/include/cudf/ast/detail/operators.hpp @@ -21,8 +21,8 @@ #include #include -#include #include +#include #include #include @@ -35,14 +35,14 @@ namespace ast { namespace detail { -// Type trait for wrapping nullable types in a cuda::std::optional. Non-nullable +// Type trait for wrapping nullable types in a thrust::optional. Non-nullable // types are returned as is. template struct possibly_null_value; template struct possibly_null_value { - using type = cuda::std::optional; + using type = thrust::optional; }; template