Skip to content
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
4 changes: 3 additions & 1 deletion cpp/src/ast/operators.cpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2021-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/
#include <cudf/ast/detail/operators.cuh>
Expand Down Expand Up @@ -151,6 +151,7 @@ struct type_dispatch_binary_op {
*
* @tparam F Type of forwarded functor.
* @tparam Ts Parameter pack of forwarded arguments.
* @param op The binary operator to dispatch
* @param lhs_type Type of left input data.
* @param rhs_type Type of right input data.
* @param f Forwarded functor to be called.
Expand Down Expand Up @@ -218,6 +219,7 @@ struct type_dispatch_unary_op {
*
* @tparam F Type of forwarded functor.
* @tparam Ts Parameter pack of forwarded arguments.
* @param op The unary operator to dispatch
* @param input_type Type of input data.
* @param f Forwarded functor to be called.
* @param args Forwarded arguments to `operator()` of `f`.
Expand Down
3 changes: 2 additions & 1 deletion cpp/src/copying/concatenate.cu
Original file line number Diff line number Diff line change
Expand Up @@ -454,7 +454,8 @@ void traverse_children::operator()<cudf::list_view>(host_span<column_view const>
* @brief Verifies that the sum of the sizes of all the columns to be concatenated
* will not exceed the max value of size_type, and verifies all column types match
*
* @param columns_to_concat Span of columns to check
* @param cols Span of columns to check
* @param stream CUDA stream used for device memory operations and kernel launches
*
* @throws cudf::logic_error if the total length of the concatenated columns would
* exceed the max value of size_type
Expand Down
10 changes: 3 additions & 7 deletions cpp/src/copying/contiguous_split.cu
Original file line number Diff line number Diff line change
Expand Up @@ -149,14 +149,8 @@ struct dst_buf_info {
* @param dst Destination buffer
* @param src Source buffer
* @param t Thread index
* @param num_elements Number of elements to copy
* @param element_size Size of each element in bytes
* @param src_element_index Element index to start copying at
* @param dst_info Destination buffer info containing element count, sizes, shifts, and validity
* @param stride Size of the kernel block
* @param value_shift Shift incoming 4-byte offset values down by this amount
* @param bit_shift Shift incoming data right by this many bits
* @param num_rows Number of rows being copied
* @param valid_count Optional pointer to a value to store count of set bits
*/
template <int block_size, bool is_offsets, typename offset_type = int32_t>
__device__ void copy_buffer(uint8_t* __restrict__ dst,
Expand Down Expand Up @@ -464,6 +458,7 @@ size_type count_src_bufs(InputIter begin, InputIter end)
* @param end End of input columns
* @param head Beginning of source buffer info array
* @param current Current source buffer info to be written to
* @param stream CUDA stream used for device memory operations and kernel launches
* @param offset_stack_pos Integer representing our current offset nesting depth
* (how many list or string levels deep we are)
* @param parent_offset_index Index into src_buf_info output array indicating our nearest
Expand Down Expand Up @@ -799,6 +794,7 @@ std::tuple<size_t, int64_t, int64_t, size_type> build_output_column_metadata(
* copied buffer
* @param out_begin Output iterator of column views
* @param base_ptr Pointer to the base address of copied data for the working partition
* @param mb Memory block for the output columns
*
* @returns new dst_buf_info iterator after processing this range of input columns
*/
Expand Down
3 changes: 2 additions & 1 deletion cpp/src/groupby/sort/group_reductions.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -304,6 +304,7 @@ std::unique_ptr<column> group_var(column_view const& values,
* @param values Grouped and sorted (within group) values to get quantiles from
* @param group_sizes Number of valid elements per group
* @param group_offsets Offsets of groups' starting points within @p values
* @param num_groups Number of groups
* @param quantiles List of quantiles q where q lies in [0,1]
* @param interp Method to use when desired value lies between data points
* @param stream CUDA stream used for device memory operations and kernel launches.
Expand Down
3 changes: 2 additions & 1 deletion cpp/src/groupby/sort/group_scan.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2021-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -169,6 +169,7 @@ std::unique_ptr<column> average_rank_scan(column_view const& grouped_values,
* @brief Internal API to calculate groupwise dense rank value
*
* @param grouped_values column or struct column that rows within a group are sorted by
* @param value_order Order of values within groups
* @param group_labels ID of group that the corresponding value belongs to
* @param group_offsets group index offsets with group ID indices
* @param stream CUDA stream used for device memory operations and kernel launches
Expand Down
8 changes: 4 additions & 4 deletions cpp/src/io/avro/avro_gpu.cu
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/
#include "avro_gpu.hpp"
Expand Down Expand Up @@ -52,7 +52,7 @@ static inline int64_t __device__ avro_decode_zigzag_varint(uint8_t const*& cur,
* destination data.
* @param[in] cur Current input data pointer
* @param[in] end End of input data
* @param[in] global_Dictionary Global dictionary entries
* @param[in] global_dictionary Global dictionary entries
* @param[out] skipped_row Whether the row was skipped; set to false
* if the row was saved (caller should ensure
* this is initialized to true)
Expand Down Expand Up @@ -305,8 +305,8 @@ avro_decode_row(schemadesc_s const* schema,
* @brief Decode column data
*
* @param[in] blocks Data block descriptions
* @param[in] schema Schema description
* @param[in] global_Dictionary Global dictionary entries
* @param[in] schema_g Schema description
* @param[in] global_dictionary Global dictionary entries
* @param[in] avro_data Raw block data
* @param[in] schema_len Number of entries in schema
* @param[in] min_row_size Minimum size in bytes of a row
Expand Down
4 changes: 1 addition & 3 deletions cpp/src/io/comp/gpuinflate.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -86,8 +86,6 @@ size_t get_gpu_debrotli_scratch_size(int max_num_inputs = 0);
* @param[in] inputs List of input buffers
* @param[out] outputs List of output buffers
* @param[out] results List of output status structures
* @param[in] scratch Temporary memory for intermediate work
* @param[in] scratch_size Size in bytes of the temporary memory
* @param[in] stream CUDA stream to use
*/
CUDF_EXPORT
Expand Down
1 change: 1 addition & 0 deletions cpp/src/io/csv/csv_gpu.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -184,6 +184,7 @@ device_span<uint64_t> remove_blank_rows(cudf::io::parse_options_view const& opti
* @param[in] data The row-column data
* @param[in] column_flags Flags that control individual column parsing
* @param[in] row_offsets List of row data start positions (offsets)
* @param[in] num_active_columns Number of active columns
* @param[in] stream CUDA stream to use
*
* @return stats Histogram of each dtypes' occurrence for each column
Expand Down
4 changes: 2 additions & 2 deletions cpp/src/io/fst/lookup_tables.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -80,7 +80,7 @@ class SingleSymbolSmemLUT {
*
* @param symbol_strings Array of strings, where the i-th string holds all symbols
* (characters!) that correspond to the i-th symbol group index
* @param stream The stream used to copy the lookup table to device
* @param pre_map_op Function object that transforms a symbol to a symbol group id
* @return
*/
template <typename SymbolGroupItT>
Expand Down Expand Up @@ -863,7 +863,7 @@ class Dfa {
* output symbols is written
* @tparam OffsetT A type large enough to index into either of both: (a) the input symbols and
* (b) the output symbols
* @param d_chars Pointer to the input string of symbols
* @param d_chars_it Pointer to the input string of symbols
* @param num_chars The total number of input symbols to process
* @param d_out_it Random-access output iterator to which the transduced output is
* written
Expand Down
2 changes: 2 additions & 0 deletions cpp/src/io/json/column_tree_construction.cu
Original file line number Diff line number Diff line change
Expand Up @@ -86,6 +86,8 @@ struct parent_nodeids_to_colids {
*
* @param node_tree Node tree representation of JSON string
* @param original_col_ids Column ids of nodes
* @param sorted_col_ids Sorted column IDs
* @param ordered_node_ids Ordered node IDs
* @param row_offsets Row offsets of nodes
* @param is_array_of_arrays Whether the tree is an array of arrays
* @param row_array_parent_col_id Column id of row array, if is_array_of_arrays is true
Expand Down
3 changes: 1 addition & 2 deletions cpp/src/io/orc/orc_gpu.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -308,7 +308,6 @@ void decode_nulls_and_string_dictionaries(column_desc* chunks,
* @param[in] num_stripes Number of stripes
* @param[in] first_row Crop all rows below first_row
* @param[in] tz_table Timezone translation table
* @param[in] tz_len Length of timezone translation table
* @param[in] row_groups Optional row index data [rowgroup][column]
* @param[in] num_rowgroups Number of row groups in row index data
* @param[in] rowidx_stride Row index stride
Expand Down
3 changes: 2 additions & 1 deletion cpp/src/io/orc/reader_impl_decode.cu
Original file line number Diff line number Diff line change
Expand Up @@ -52,6 +52,7 @@ namespace {
* @param loaded_stripe_range Range of stripes that are already loaded in memory
* @param stream_range Range of streams to be decoded
* @param num_decode_stripes Number of stripes that the decoding streams belong to
* @param compinfo Compression information for each stripe
* @param compinfo_map A map to lookup compression info of streams
* @param decompressor Block decompressor
* @param stripe_data List of source stripe column data
Expand Down Expand Up @@ -346,7 +347,7 @@ void update_null_mask(cudf::detail::hostdevice_2dvector<column_desc>& chunks,
* @param skip_rows Number of rows to offset from start
* @param row_index_stride Distance between each row index
* @param level Current nesting level being processed
* @param tz_table Local time to UTC conversion table
* @param d_tz_table Local time to UTC conversion table
* @param chunks Vector of list of column chunk descriptors
* @param row_groups Vector of list of row index descriptors
* @param out_buffers Output columns' device buffers
Expand Down
3 changes: 2 additions & 1 deletion cpp/src/io/orc/stats_enc.cu
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -477,6 +477,7 @@ void orc_init_statistics_buffersize(statistics_merge_group* groups,
* @param[in,out] groups Statistics merge groups
* @param[in,out] chunks Statistics data
* @param[in] statistics_count Number of statistics buffers
* @param[in] stream CUDA stream used for device memory operations and kernel launches
*/
void orc_encode_statistics(uint8_t* blob_bfr,
statistics_merge_group* groups,
Expand Down
10 changes: 6 additions & 4 deletions cpp/src/io/orc/stripe_data.cu
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -378,7 +378,6 @@ inline __device__ uint32_t bytestream_readu32(orc_bytestream_s* bs, int pos)
*
* @param[in] bs Byte stream input
* @param[in] pos Position in byte stream
* @param[in] numbits number of bits
* @return bits
*/
inline __device__ uint64_t bytestream_readu64(orc_bytestream_s* bs, int pos)
Expand Down Expand Up @@ -1165,10 +1164,12 @@ static const __device__ __constant__ int64_t kPow5i[28] = {1,
* @brief ORC Decimal decoding (unbounded base-128 varints)
*
* @param[in] bs Input byte stream
* @param[in] scratch Scratch buffer for intermediate results
* @param[in,out] vals on input: scale from secondary stream, on output: value
* @param[in] val_scale Scale of each value
* @param[in] col_scale Scale from schema to which value will be adjusted
* @param[in] numvals Number of values to decode
* @param[in] dtype_id Data type identifier for the decimal column
* @param[in] col_scale Scale from schema to which value will be adjusted
* @param[in] t thread id
*
* @return number of values decoded
Expand Down Expand Up @@ -2071,14 +2072,15 @@ void __host__ decode_nulls_and_string_dictionaries(column_desc* chunks,
*
* @param[in] chunks column_desc device array [stripe][column]
* @param[in] global_dictionary Global dictionary device array
* @param[in] row_groups Optional row index data [row_group][column]
* @param[in] num_columns Number of columns
* @param[in] num_stripes Number of stripes
* @param[in] first_row Crop all rows below first_row
* @param[in] tz_table Timezone translation table
* @param[in] row_groups Optional row index data [row_group][column]
* @param[in] num_rowgroups Number of row groups in row index data
* @param[in] rowidx_stride Row index stride
* @param[in] level nesting level being processed
* @param[in] error_count Pointer to error count output
* @param[in] stream CUDA stream used for device memory operations and kernel launches
*/
void __host__ decode_column_data(column_desc* chunks,
Expand Down
3 changes: 1 addition & 2 deletions cpp/src/io/orc/stripe_enc.cu
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -340,7 +340,6 @@ static inline __device__ void StoreBitsBigEndian(
* @param[in] inbuf base input buffer
* @param[in] inpos position in input buffer
* @param[in] numvals max number of values to encode
* @param[in] flush encode all remaining values if nonzero
* @param[in] t thread id
* @param[in] temp_storage shared memory storage to perform block reduce
*
Expand Down
10 changes: 7 additions & 3 deletions cpp/src/io/orc/writer_impl.cu
Original file line number Diff line number Diff line change
Expand Up @@ -442,6 +442,7 @@ namespace {
* @param columns List of columns
* @param rowgroup_bounds Ranges of rows in each rowgroup [rowgroup][column]
* @param max_stripe_size Maximum size of each stripe, both in bytes and in rows
* @param stream CUDA stream used for device memory operations and kernel launches
* @return List of stripe descriptors
*/
file_segmentation calculate_segmentation(host_span<orc_column_view const> columns,
Expand Down Expand Up @@ -536,6 +537,9 @@ size_t rle_stream_size(TypeKind kind, size_t count)
* @param[in,out] columns List of columns
* @param[in] segmentation stripe and rowgroup ranges
* @param[in] decimal_column_sizes Sizes of encoded decimal columns
* @param[in] enable_dictionary Whether dictionary encoding is enabled
* @param[in] compression Compression type to use
* @param[in] write_mode The write mode (single or chunked)
* @return List of stream descriptors
*/
orc_streams create_streams(host_span<orc_column_view> columns,
Expand Down Expand Up @@ -1202,7 +1206,7 @@ cudf::detail::hostdevice_vector<uint8_t> allocate_and_encode_blobs(
/**
* @brief Returns column statistics in an intermediate format.
*
* @param statistics_freq Frequency of statistics to be included in the output file
* @param stats_freq Frequency of statistics to be included in the output file
* @param orc_table Table information to be written
* @param segmentation stripe and rowgroup ranges
* @param stream CUDA stream used for device memory operations and kernel launches
Expand Down Expand Up @@ -1326,8 +1330,8 @@ intermediate_statistics gather_statistic_blobs(statistics_freq const stats_freq,
/**
* @brief Returns column statistics encoded in ORC protobuf format stored in the footer.
*
* @param num_stripes number of stripes in the data
* @param incoming_stats intermediate statistics returned from `gather_statistic_blobs`
* @param footer ORC footer containing stripe and type information
* @param per_chunk_stats Persisted per-chunk statistics from `gather_statistic_blobs`
* @param stream CUDA stream used for device memory operations and kernel launches
* @return The encoded statistic blobs
*/
Expand Down
7 changes: 4 additions & 3 deletions cpp/src/io/parquet/decode_fixed.cu
Original file line number Diff line number Diff line change
Expand Up @@ -272,11 +272,11 @@ __device__ inline void decode_fixed_width_split_values(
*
* @tparam decode_block_size Size of the thread block
* @tparam level_t Definition level type
* @tparam is_nested Whether the type is nested
*
* @param target_value_count The target value count to process
* @param s Pointer to page state
* @param def Pointer to the definition levels
* @param is_nested Whether the column is nested
* @param t Thread index
*
* @return Maximum depth valid count after skipping
Expand Down Expand Up @@ -548,9 +548,10 @@ __device__ int update_validity_and_row_indices_flat(
* @tparam state_buf State buffer type
*
* @param target_value_count The target value count to process
* @param s Pointer to page state
* @param sb Pointer to state buffer
* @param s Pointer to page state
* @param sb Pointer to state buffer
* @param def Pointer to the definition levels
* @param rep Repetition level decoder
* @param t Thread index
*
* @return Maximum depth valid count after processing
Expand Down
1 change: 1 addition & 0 deletions cpp/src/io/parquet/decode_preprocess.cu
Original file line number Diff line number Diff line change
Expand Up @@ -42,6 +42,7 @@ using unused_state_buf = page_state_buffers_s<0, 0, 0>;
* @param rep Repetition level buffer
* @param def Definition level buffer
* @param bounds_set Boolean indicating whether min/max row bounds have been set
* @param block The cooperative thread block
*/
template <typename level_t>
__device__ void update_page_sizes(page_state_s* s,
Expand Down
4 changes: 2 additions & 2 deletions cpp/src/io/parquet/error.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* SPDX-FileCopyrightText: Copyright (c) 2023-2025, NVIDIA CORPORATION.
* SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION.
* SPDX-License-Identifier: Apache-2.0
*/

Expand Down Expand Up @@ -38,7 +38,7 @@ class kernel_error {
* @throws `rmm::bad_alloc` if allocating the device memory for `initial_value` fails.
* @throws `rmm::cuda_error` if copying `initial_value` to device memory fails.
*
* @param CUDA stream to use
* @param stream CUDA stream to use
*/
kernel_error(rmm::cuda_stream_view stream) : _error_code(1, stream)
{
Expand Down
4 changes: 2 additions & 2 deletions cpp/src/io/parquet/experimental/deletion_vectors.cu
Original file line number Diff line number Diff line change
Expand Up @@ -337,8 +337,8 @@ std::unique_ptr<cudf::column> compute_row_mask_column(
* vectors
*
* @param row_index_column View of the row index column
* @param deletion_vector_refs Queue of roaring bitmap wrappers
* @param deletion_vector_row_counts Queue of number of rows in eachdeletion vector
* @param deletion_vectors Queue of roaring bitmap wrappers
Comment thread
vuule marked this conversation as resolved.
* @param deletion_vector_row_counts Queue of number of rows in each deletion vector
* @param start_row Starting row index of the current table chunk
* @param stream CUDA stream for kernel launches and data transfers
* @param mr Device memory resource to allocate device memory for the row mask column
Expand Down
2 changes: 2 additions & 0 deletions cpp/src/io/parquet/experimental/dictionary_page_filter.cu
Original file line number Diff line number Diff line change
Expand Up @@ -1007,6 +1007,8 @@ struct dictionary_caster {
* @brief Build BOOL8 columns from dictionary membership results device buffers
*
* @param results_buffers Vector of dictionary membership results device buffers
* @param stream CUDA stream used for device memory operations and kernel launches
* @param mr Device memory resource used to allocate the returned device memory
*
* @return A vector of BOOL8 columns
*/
Expand Down
Loading
Loading