diff --git a/cub/cub/block/block_discontinuity.cuh b/cub/cub/block/block_discontinuity.cuh index 2fb15e9059b..fb88dfac07f 100644 --- a/cub/cub/block/block_discontinuity.cuh +++ b/cub/cub/block/block_discontinuity.cuh @@ -270,7 +270,7 @@ public: //! @name Head flag operations //! @{ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * @param[out] head_flags @@ -349,7 +349,7 @@ public: Iterate::FlagHeads(linear_tid, head_flags, input, preds, flag_op); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sets head flags indicating discontinuities between items partitioned across the thread diff --git a/cub/cub/block/block_exchange.cuh b/cub/cub/block/block_exchange.cuh index a781d68e68b..bdc2a3dc932 100644 --- a/cub/cub/block/block_exchange.cuh +++ b/cub/cub/block/block_exchange.cuh @@ -1217,7 +1217,7 @@ public: //! @} end member group -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// @param[in-out] items /// Items to exchange, converting between **striped** and **blocked** arrangements. @@ -1292,7 +1292,7 @@ public: ScatterToStriped(items, items, ranks, is_valid); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/block/block_load.cuh b/cub/cub/block/block_load.cuh index 641ff6d5d09..c1e9b95ac56 100644 --- a/cub/cub/block/block_load.cuh +++ b/cub/cub/block/block_load.cuh @@ -179,7 +179,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE void LoadDirectBlocked( LoadDirectBlocked(linear_tid, block_src_it, dst_items, block_items_end); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document //! @brief Internal implementation for load vectorization //! @@ -225,7 +225,7 @@ InternalLoadDirectBlockedVectorized(int linear_tid, const T* block_src_ptr, T (& } } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Load a linear segment of items into a blocked arrangement across the thread block. diff --git a/cub/cub/block/block_merge_sort.cuh b/cub/cub/block/block_merge_sort.cuh index 29510db5e97..b6d0c8a33b1 100644 --- a/cub/cub/block/block_merge_sort.cuh +++ b/cub/cub/block/block_merge_sort.cuh @@ -175,14 +175,14 @@ private: // Whether or not there are values to be trucked along with keys static constexpr bool KEYS_ONLY = ::cuda::std::is_same::value; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Shared memory type required by this thread block union _TempStorage { KeyT keys_shared[ITEMS_PER_TILE + 1]; ValueT items_shared[ITEMS_PER_TILE + 1]; }; // union TempStorage -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /// Shared storage reference _TempStorage& temp_storage; diff --git a/cub/cub/block/block_radix_rank.cuh b/cub/cub/block/block_radix_rank.cuh index 73228368fc5..5426e967712 100644 --- a/cub/cub/block/block_radix_rank.cuh +++ b/cub/cub/block/block_radix_rank.cuh @@ -93,7 +93,7 @@ struct BlockRadixRankEmptyCallback _CCCL_DEVICE _CCCL_FORCEINLINE void operator()(int (&bins)[BINS_PER_THREAD]) {} }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -121,7 +121,7 @@ struct warp_in_block_matcher_t }; } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! BlockRadixRank provides operations for ranking unsigned integer types within a CUDA thread block. @@ -263,7 +263,7 @@ private: /// BlockScan type using BlockScan = BlockScan; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document struct __align__(16) _TempStorage { union Aliasable @@ -276,7 +276,7 @@ private: // Storage for scanning local ranks typename BlockScan::TempStorage block_scan; }; -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED /// Shared storage reference _TempStorage& temp_storage; @@ -597,7 +597,7 @@ private: /// BlockScan type using BlockScanT = BlockScan; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document struct __align__(16) _TempStorage { typename BlockScanT::TempStorage block_scan; @@ -609,7 +609,7 @@ private: } aliasable; }; -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED /// Shared storage reference _TempStorage& temp_storage; @@ -1183,7 +1183,7 @@ struct BlockRadixRankMatchEarlyCounts } }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -1211,6 +1211,6 @@ using block_radix_rank_t = ::cuda::std::_If< BlockRadixRankMatchEarlyCounts>>>>; } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/block/block_radix_sort.cuh b/cub/cub/block/block_radix_sort.cuh index 48650992918..3223b920b13 100644 --- a/cub/cub/block/block_radix_sort.cuh +++ b/cub/cub/block/block_radix_sort.cuh @@ -303,7 +303,7 @@ private: /// BlockExchange utility type for values using BlockExchangeValues = BlockExchange; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Shared memory storage layout type union _TempStorage { @@ -312,7 +312,7 @@ private: typename BlockExchangeKeys::TempStorage exchange_keys; typename BlockExchangeValues::TempStorage exchange_values; }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /****************************************************************************** * Thread fields @@ -469,7 +469,7 @@ private: } public: -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * @brief Sort blocked -> striped arrangement @@ -554,7 +554,7 @@ public: } } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /// @smemstorage{BlockRadixSort} struct TempStorage : Uninitialized<_TempStorage> diff --git a/cub/cub/block/block_run_length_decode.cuh b/cub/cub/block/block_run_length_decode.cuh index 253fdb8b1d9..74934576cd5 100644 --- a/cub/cub/block/block_run_length_decode.cuh +++ b/cub/cub/block/block_run_length_decode.cuh @@ -173,7 +173,7 @@ private: /// Type used to index into the block's runs using RunOffsetT = uint32_t; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Shared memory type required by this thread block union _TempStorage { @@ -184,7 +184,7 @@ private: DecodedOffsetT run_offsets[BLOCK_RUNS]; } runs; }; // union TempStorage -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /// Internal storage allocator (used when the user does not provide pre-allocated shared memory) _CCCL_DEVICE _CCCL_FORCEINLINE _TempStorage& PrivateStorage() diff --git a/cub/cub/block/block_scan.cuh b/cub/cub/block/block_scan.cuh index 0644e8ca254..c49eb36a52e 100644 --- a/cub/cub/block/block_scan.cuh +++ b/cub/cub/block/block_scan.cuh @@ -1291,7 +1291,7 @@ public: } //! @} end member group -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document no-initial-value scans +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document no-initial-value scans //! @name Exclusive prefix scan operations (no initial value, single datum per thread) //! @{ @@ -1445,7 +1445,7 @@ public: } //! @} end member group -#endif // DOXYGEN_SHOULD_SKIP_THIS // Do not document no-initial-value scans +#endif // _CCCL_DOXYGEN_INVOKED // Do not document no-initial-value scans //! @name Inclusive prefix sum operations //! @{ diff --git a/cub/cub/block/block_store.cuh b/cub/cub/block/block_store.cuh index 9d057d7fe4b..443f7a7f93b 100644 --- a/cub/cub/block/block_store.cuh +++ b/cub/cub/block/block_store.cuh @@ -1229,12 +1229,12 @@ public: //! @} end member group }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template > struct BlockStoreType { using type = cub::BlockStore; }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/block/radix_rank_sort_operations.cuh b/cub/cub/block/radix_rank_sort_operations.cuh index e56a0ec1e27..d4fdd9c405f 100644 --- a/cub/cub/block/radix_rank_sort_operations.cuh +++ b/cub/cub/block/radix_rank_sort_operations.cuh @@ -142,7 +142,7 @@ struct ShiftDigitExtractor : BaseDigitExtractor } }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -564,7 +564,7 @@ struct traits_t } // namespace radix } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! Twiddling keys for radix sort template diff --git a/cub/cub/detail/array_utils.cuh b/cub/cub/detail/array_utils.cuh index cfc8fafb452..1857c895a3c 100644 --- a/cub/cub/detail/array_utils.cuh +++ b/cub/cub/detail/array_utils.cuh @@ -51,7 +51,7 @@ CUB_NAMESPACE_BEGIN namespace detail { -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /*********************************************************************************************************************** * Generic Array-like to Array Conversion @@ -74,7 +74,7 @@ to_array(const Input& input) return to_array_impl(input, ::cuda::std::make_index_sequence()>{}); } -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED } // namespace detail diff --git a/cub/cub/detail/detect_cuda_runtime.cuh b/cub/cub/detail/detect_cuda_runtime.cuh index 35c52f4aedb..d83b2c1179a 100644 --- a/cub/cub/detail/detect_cuda_runtime.cuh +++ b/cub/cub/detail/detect_cuda_runtime.cuh @@ -49,7 +49,7 @@ # include #endif // !_CCCL_COMPILER(NVRTC) -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Only parse this during doxygen passes: +#ifdef _CCCL_DOXYGEN_INVOKED // Only parse this during doxygen passes: /** * \def CUB_DISABLE_CDP diff --git a/cub/cub/detail/nvtx.cuh b/cub/cub/detail/nvtx.cuh index 6a5dd8ff039..3bda5e596f3 100644 --- a/cub/cub/detail/nvtx.cuh +++ b/cub/cub/detail/nvtx.cuh @@ -37,10 +37,10 @@ # pragma system_header #endif // no system header -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Only parse this during doxygen passes: +#ifdef _CCCL_DOXYGEN_INVOKED // Only parse this during doxygen passes: //! When this macro is defined, no NVTX ranges are emitted by CCCL # define CCCL_DISABLE_NVTX -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED // Enable the functionality of this header if: // * The NVTX3 C API is available in CTK diff --git a/cub/cub/detail/strong_load.cuh b/cub/cub/detail/strong_load.cuh index e63bd3456c0..61693d808e2 100644 --- a/cub/cub/detail/strong_load.cuh +++ b/cub/cub/detail/strong_load.cuh @@ -49,7 +49,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -247,6 +247,6 @@ static _CCCL_DEVICE _CCCL_FORCEINLINE unsigned int load_acquire(unsigned int con } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/detail/strong_store.cuh b/cub/cub/detail/strong_store.cuh index fe16cae9674..9b8091738db 100644 --- a/cub/cub/detail/strong_store.cuh +++ b/cub/cub/detail/strong_store.cuh @@ -47,7 +47,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -302,6 +302,6 @@ _CCCL_DEVICE _CCCL_FORCEINLINE void store_release(unsigned char* ptr, unsigned c } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/device/device_adjacent_difference.cuh b/cub/cub/device/device_adjacent_difference.cuh index 84add4262e2..41728342abc 100644 --- a/cub/cub/device/device_adjacent_difference.cuh +++ b/cub/cub/device/device_adjacent_difference.cuh @@ -267,7 +267,7 @@ public: d_temp_storage, temp_storage_bytes, d_input, d_output, num_items, difference_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED static CUB_RUNTIME_FUNCTION cudaError_t SubtractLeftCopy( void* d_temp_storage, @@ -283,7 +283,7 @@ public: return SubtractLeftCopy(d_temp_storage, temp_storage_bytes, d_input, d_output, num_items, difference_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Subtracts the left element of each adjacent pair of elements residing within device-accessible memory. @@ -398,7 +398,7 @@ public: d_temp_storage, temp_storage_bytes, d_input, d_input, num_items, difference_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED static CUB_RUNTIME_FUNCTION cudaError_t SubtractLeft( void* d_temp_storage, @@ -413,7 +413,7 @@ public: return SubtractLeft(d_temp_storage, temp_storage_bytes, d_input, num_items, difference_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Subtracts the right element of each adjacent pair of elements residing within device-accessible memory. @@ -545,7 +545,7 @@ public: d_temp_storage, temp_storage_bytes, d_input, d_output, num_items, difference_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED static CUB_RUNTIME_FUNCTION cudaError_t SubtractRightCopy( void* d_temp_storage, @@ -561,7 +561,7 @@ public: return SubtractRightCopy(d_temp_storage, temp_storage_bytes, d_input, d_output, num_items, difference_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Subtracts the right element of each adjacent pair of elements residing within device-accessible memory. @@ -665,7 +665,7 @@ public: d_temp_storage, temp_storage_bytes, d_input, d_input, num_items, difference_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED static CUB_RUNTIME_FUNCTION cudaError_t SubtractRight( void* d_temp_storage, @@ -680,7 +680,7 @@ public: return SubtractRight(d_temp_storage, temp_storage_bytes, d_input, num_items, difference_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/device_histogram.cuh b/cub/cub/device/device_histogram.cuh index e6abc4bd07b..32e485df2b3 100644 --- a/cub/cub/device/device_histogram.cuh +++ b/cub/cub/device/device_histogram.cuh @@ -206,7 +206,7 @@ struct DeviceHistogram stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t HistogramEven( void* d_temp_storage, @@ -233,7 +233,7 @@ struct DeviceHistogram num_samples, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes an intensity histogram from a sequence of data samples using equal-width bins. @@ -386,7 +386,7 @@ struct DeviceHistogram stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t HistogramEven( void* d_temp_storage, @@ -417,7 +417,7 @@ struct DeviceHistogram row_stride_bytes, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes per-channel intensity histograms from a sequence of multi-channel "pixel" data samples using @@ -588,7 +588,7 @@ struct DeviceHistogram stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t HistogramRange( void* d_temp_storage, @@ -1017,7 +1017,7 @@ struct DeviceHistogram return HistogramRange( d_temp_storage, temp_storage_bytes, d_samples, d_histogram, num_levels, d_levels, num_samples, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes an intensity histogram from a sequence of data samples using the specified bin boundary levels. @@ -1157,7 +1157,7 @@ struct DeviceHistogram stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t HistogramRange( void* d_temp_storage, @@ -1186,7 +1186,7 @@ struct DeviceHistogram row_stride_bytes, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes per-channel intensity histograms from a sequence of multi-channel "pixel" data samples @@ -1346,7 +1346,7 @@ struct DeviceHistogram stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -263,7 +263,7 @@ public: return SortPairs( d_temp_storage, temp_storage_bytes, d_keys, d_items, num_items, compare_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Sorts items using a merge sorting method. @@ -411,7 +411,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -586,7 +586,7 @@ public: return SortKeys( d_temp_storage, temp_storage_bytes, d_keys, num_items, compare_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -729,7 +729,7 @@ public: d_temp_storage, temp_storage_bytes, d_input_keys, d_output_keys, num_items, compare_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysCopy( void* d_temp_storage, @@ -746,7 +746,7 @@ public: return SortKeysCopy( d_temp_storage, temp_storage_bytes, d_input_keys, d_output_keys, num_items, compare_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Sorts items using a merge sorting method. @@ -857,7 +857,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_items, num_items, compare_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortPairs( void* d_temp_storage, @@ -874,7 +874,7 @@ public: return StableSortPairs( d_temp_storage, temp_storage_bytes, d_keys, d_items, num_items, compare_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Sorts items using a merge sorting method. @@ -976,7 +976,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, num_items, compare_op, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortKeys( void* d_temp_storage, @@ -992,7 +992,7 @@ public: return StableSortKeys( d_temp_storage, temp_storage_bytes, d_keys, num_items, compare_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Sorts items using a merge sorting method. diff --git a/cub/cub/device/device_partition.cuh b/cub/cub/device/device_partition.cuh index 48666f1370b..621bf2b9070 100644 --- a/cub/cub/device/device_partition.cuh +++ b/cub/cub/device/device_partition.cuh @@ -223,7 +223,7 @@ struct DevicePartition stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_in, d_flags, d_out, d_num_selected_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Uses the ``select_op`` functor to split the corresponding items from ``d_in`` into @@ -405,7 +405,7 @@ struct DevicePartition stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_in, d_out, d_num_selected_out, num_items, select_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -818,7 +818,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -1252,7 +1252,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -1706,7 +1706,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -2412,7 +2412,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -2552,7 +2552,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -2945,7 +2945,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, @@ -3345,7 +3345,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, d_values, num_items, begin_bit, end_bit, is_overwrite_okay, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, diff --git a/cub/cub/device/device_reduce.cuh b/cub/cub/device/device_reduce.cuh index a9b94f60534..bd78224be5d 100644 --- a/cub/cub/device/device_reduce.cuh +++ b/cub/cub/device/device_reduce.cuh @@ -205,7 +205,7 @@ struct DeviceReduce d_temp_storage, temp_storage_bytes, d_in, d_out, static_cast(num_items), reduction_op, init, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Reduce( void* d_temp_storage, @@ -223,7 +223,7 @@ struct DeviceReduce return Reduce( d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, reduction_op, init, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide sum using the addition (``+``) operator. @@ -330,7 +330,7 @@ struct DeviceReduce stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Sum(void* d_temp_storage, @@ -345,7 +345,7 @@ struct DeviceReduce return Sum(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide minimum using the less-than (``<``) operator. @@ -456,7 +456,7 @@ struct DeviceReduce stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Min(void* d_temp_storage, @@ -471,7 +471,7 @@ struct DeviceReduce return Min(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Finds the first device-wide minimum using the less-than (``<``) operator, also returning the index of that item. @@ -591,7 +591,7 @@ struct DeviceReduce d_temp_storage, temp_storage_bytes, d_indexed_in, d_out, num_items, cub::ArgMin(), initial_value, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ArgMin( void* d_temp_storage, @@ -606,7 +606,7 @@ struct DeviceReduce return ArgMin(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide maximum using the greater-than (``>``) operator. @@ -715,7 +715,7 @@ struct DeviceReduce stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Max(void* d_temp_storage, @@ -730,7 +730,7 @@ struct DeviceReduce return Max(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Finds the first device-wide maximum using the greater-than (``>``) @@ -854,7 +854,7 @@ struct DeviceReduce d_temp_storage, temp_storage_bytes, d_indexed_in, d_out, num_items, cub::ArgMax(), initial_value, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ArgMax( void* d_temp_storage, @@ -869,7 +869,7 @@ struct DeviceReduce return ArgMax(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Fuses transform and reduce operations @@ -1195,7 +1195,7 @@ struct DeviceReduce stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_in, d_unique_out, d_counts_out, d_num_runs_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Enumerates the starting offsets and lengths of all non-trivial runs @@ -386,7 +386,7 @@ struct DeviceRunLengthEncode stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_in, d_offsets_out, d_lengths_out, d_num_runs_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/device_scan.cuh b/cub/cub/device/device_scan.cuh index e8b56709eda..e105fa36819 100644 --- a/cub/cub/device/device_scan.cuh +++ b/cub/cub/device/device_scan.cuh @@ -208,7 +208,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ExclusiveSum( void* d_temp_storage, @@ -224,7 +224,7 @@ struct DeviceScan return ExclusiveSum( d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix sum in-place. @@ -302,7 +302,7 @@ struct DeviceScan return ExclusiveSum(d_temp_storage, temp_storage_bytes, d_data, d_data, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ExclusiveSum( void* d_temp_storage, @@ -316,7 +316,7 @@ struct DeviceScan return ExclusiveSum(d_temp_storage, temp_storage_bytes, d_data, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix scan using the specified @@ -450,7 +450,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ExclusiveScan( void* d_temp_storage, @@ -468,7 +468,7 @@ struct DeviceScan return ExclusiveScan( d_temp_storage, temp_storage_bytes, d_in, d_out, scan_op, init_value, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix scan using the specified @@ -579,7 +579,7 @@ struct DeviceScan return ExclusiveScan(d_temp_storage, temp_storage_bytes, d_data, d_data, scan_op, init_value, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ExclusiveScan( void* d_temp_storage, @@ -596,7 +596,7 @@ struct DeviceScan return ExclusiveScan( d_temp_storage, temp_storage_bytes, d_data, scan_op, init_value, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix scan using the specified @@ -739,7 +739,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_in, d_out, scan_op, init_value, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix scan using the specified binary ``scan_op`` functor. @@ -880,7 +880,7 @@ struct DeviceScan return ExclusiveScan(d_temp_storage, temp_storage_bytes, d_data, d_data, scan_op, init_value, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_data, scan_op, init_value, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group //! @name Inclusive scans @@ -1003,7 +1003,7 @@ struct DeviceScan d_temp_storage, temp_storage_bytes, d_in, d_out, ::cuda::std::plus<>{}, NullType{}, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t InclusiveSum( void* d_temp_storage, @@ -1019,7 +1019,7 @@ struct DeviceScan return InclusiveSum( d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide inclusive prefix sum in-place. @@ -1096,7 +1096,7 @@ struct DeviceScan return InclusiveSum(d_temp_storage, temp_storage_bytes, d_data, d_data, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t InclusiveSum( void* d_temp_storage, @@ -1110,7 +1110,7 @@ struct DeviceScan return InclusiveSum(d_temp_storage, temp_storage_bytes, d_data, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide inclusive prefix scan using the specified binary ``scan_op`` functor. @@ -1333,7 +1333,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t InclusiveScan( void* d_temp_storage, @@ -1350,7 +1350,7 @@ struct DeviceScan return InclusiveScan( d_temp_storage, temp_storage_bytes, d_in, d_out, scan_op, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide inclusive prefix scan using the specified binary ``scan_op`` functor. @@ -1451,7 +1451,7 @@ struct DeviceScan return InclusiveScan(d_temp_storage, temp_storage_bytes, d_data, d_data, scan_op, num_items, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t InclusiveScan( void* d_temp_storage, @@ -1466,7 +1466,7 @@ struct DeviceScan return InclusiveScan(d_temp_storage, temp_storage_bytes, d_data, scan_op, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix sum-by-key with key equality @@ -1608,7 +1608,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_keys_in, d_values_in, d_values_out, num_items, equality_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide exclusive prefix scan-by-key using the @@ -1814,7 +1814,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_keys_in, d_values_in, d_values_out, num_items, equality_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide inclusive prefix scan-by-key using the @@ -2180,7 +2180,7 @@ struct DeviceScan stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ( d_temp_storage, temp_storage_bytes, d_keys_in, d_values_in, d_values_out, scan_op, num_items, equality_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group }; diff --git a/cub/cub/device/device_segmented_radix_sort.cuh b/cub/cub/device/device_segmented_radix_sort.cuh index cc627b971ca..490caf36c48 100644 --- a/cub/cub/device/device_segmented_radix_sort.cuh +++ b/cub/cub/device/device_segmented_radix_sort.cuh @@ -265,7 +265,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -300,7 +300,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into ascending order. (``~N`` auxiliary storage required) @@ -476,7 +476,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -507,7 +507,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into descending order. (``~2N`` auxiliary storage required). @@ -683,7 +683,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -718,7 +718,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into descending order. (``~N`` auxiliary storage required). @@ -898,7 +898,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -929,7 +929,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group //! @name Keys-only @@ -1092,7 +1092,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -1123,7 +1123,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into ascending order. (``~N`` auxiliary storage required). @@ -1291,7 +1291,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -1320,7 +1320,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into descending order. (``~2N`` auxiliary storage required). @@ -1479,7 +1479,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, @@ -1510,7 +1510,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into descending order. (``~N`` auxiliary storage required). @@ -1676,7 +1676,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, @@ -1705,7 +1705,7 @@ public: end_bit, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group }; diff --git a/cub/cub/device/device_segmented_reduce.cuh b/cub/cub/device/device_segmented_reduce.cuh index 9d4de803e86..7ad043eab5f 100644 --- a/cub/cub/device/device_segmented_reduce.cuh +++ b/cub/cub/device/device_segmented_reduce.cuh @@ -272,7 +272,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Sum(void* d_temp_storage, @@ -444,7 +444,7 @@ public: return Sum( d_temp_storage, temp_storage_bytes, d_in, d_out, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide segmented minimum using the less-than (``<``) operator. @@ -572,7 +572,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Min(void* d_temp_storage, @@ -590,7 +590,7 @@ public: return Min( d_temp_storage, temp_storage_bytes, d_in, d_out, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Finds the first device-wide minimum in each segment using the @@ -742,7 +742,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ArgMin( void* d_temp_storage, @@ -760,7 +760,7 @@ public: return ArgMin( d_temp_storage, temp_storage_bytes, d_in, d_out, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Computes a device-wide segmented maximum using the greater-than (``>``) operator. @@ -877,7 +877,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Max(void* d_temp_storage, @@ -895,7 +895,7 @@ public: return Max( d_temp_storage, temp_storage_bytes, d_in, d_out, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Finds the first device-wide maximum in each segment using the @@ -1050,7 +1050,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t ArgMax( void* d_temp_storage, @@ -1068,7 +1068,7 @@ public: return ArgMax( d_temp_storage, temp_storage_bytes, d_in, d_out, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/device_segmented_sort.cuh b/cub/cub/device/device_segmented_sort.cuh index 1f219aebd25..10b5c6d2388 100644 --- a/cub/cub/device/device_segmented_sort.cuh +++ b/cub/cub/device/device_segmented_sort.cuh @@ -306,7 +306,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -333,7 +333,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -503,7 +503,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, @@ -530,7 +530,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -702,7 +702,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeys( void* d_temp_storage, @@ -720,7 +720,7 @@ public: return SortKeys( d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -893,7 +893,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortKeysDescending( void* d_temp_storage, @@ -911,7 +911,7 @@ public: return SortKeysDescending( d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into ascending order. Approximately @@ -1049,7 +1049,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortKeys( void* d_temp_storage, @@ -1076,7 +1076,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into descending order. @@ -1214,7 +1214,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortKeysDescending( void* d_temp_storage, @@ -1241,7 +1241,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into ascending order. @@ -1381,7 +1381,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortKeys( void* d_temp_storage, @@ -1399,7 +1399,7 @@ public: return StableSortKeys( d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of keys into descending order. @@ -1538,7 +1538,7 @@ public: d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortKeysDescending( void* d_temp_storage, @@ -1556,7 +1556,7 @@ public: return StableSortKeysDescending( d_temp_storage, temp_storage_bytes, d_keys, num_items, num_segments, d_begin_offsets, d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -1757,7 +1757,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -1788,7 +1788,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -1985,7 +1985,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -2016,7 +2016,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -2213,7 +2213,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairs( void* d_temp_storage, @@ -2240,7 +2240,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // Internal version without NVTX range @@ -2436,7 +2436,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t SortPairsDescending( void* d_temp_storage, @@ -2463,7 +2463,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into ascending order. @@ -2623,7 +2623,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortPairs( void* d_temp_storage, @@ -2654,7 +2654,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into descending order. @@ -2814,7 +2814,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortPairsDescending( void* d_temp_storage, @@ -2845,7 +2845,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into ascending order. @@ -3011,7 +3011,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortPairs( void* d_temp_storage, @@ -3038,7 +3038,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Sorts segments of key-value pairs into descending order. @@ -3203,7 +3203,7 @@ public: stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t StableSortPairsDescending( void* d_temp_storage, @@ -3230,7 +3230,7 @@ public: d_end_offsets, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group }; diff --git a/cub/cub/device/device_select.cuh b/cub/cub/device/device_select.cuh index b537ab9204b..27a18cf809a 100644 --- a/cub/cub/device/device_select.cuh +++ b/cub/cub/device/device_select.cuh @@ -203,7 +203,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Flagged( void* d_temp_storage, @@ -221,7 +221,7 @@ struct DeviceSelect return Flagged( d_temp_storage, temp_storage_bytes, d_in, d_flags, d_out, d_num_selected_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Uses the ``d_flags`` sequence to selectively compact the items in `d_data``. @@ -341,7 +341,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Flagged( void* d_temp_storage, @@ -358,7 +358,7 @@ struct DeviceSelect return Flagged( d_temp_storage, temp_storage_bytes, d_data, d_flags, d_num_selected_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Uses the ``select_op`` functor to selectively copy items from ``d_in`` into ``d_out``. @@ -498,7 +498,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t If(void* d_temp_storage, @@ -516,7 +516,7 @@ struct DeviceSelect return If( d_temp_storage, temp_storage_bytes, d_in, d_out, d_num_selected_out, num_items, select_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Uses the ``select_op`` functor to selectively compact items in ``d_data``. @@ -648,7 +648,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t If(void* d_temp_storage, @@ -665,7 +665,7 @@ struct DeviceSelect return If( d_temp_storage, temp_storage_bytes, d_data, d_num_selected_out, num_items, select_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Uses the ``select_op`` functor applied to ``d_flags`` to selectively copy the @@ -1011,7 +1011,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Unique( void* d_temp_storage, @@ -1028,7 +1028,7 @@ struct DeviceSelect return Unique( d_temp_storage, temp_storage_bytes, d_in, d_out, d_num_selected_out, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Given an input sequence ``d_keys_in`` and ``d_values_in`` with runs of key-value pairs with consecutive @@ -1330,7 +1330,7 @@ struct DeviceSelect stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template ::Dispatch(d_temp_storage, temp_storage_bytes, spmv_params, stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t CsrMV( void* d_temp_storage, @@ -239,7 +239,7 @@ struct DeviceSpmv num_nonzeros, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @} end member group }; diff --git a/cub/cub/device/device_transform.cuh b/cub/cub/device/device_transform.cuh index 984109692f6..ef00248b448 100644 --- a/cub/cub/device/device_transform.cuh +++ b/cub/cub/device/device_transform.cuh @@ -66,7 +66,7 @@ struct DeviceTransform ::cuda::std::move(inputs), ::cuda::std::move(output), num_items, ::cuda::std::move(transform_op), stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // This overload has additional parameters to specify temporary storage. Provided for compatibility with other CUB // APIs. template @@ -88,7 +88,7 @@ struct DeviceTransform return Transform( ::cuda::std::move(inputs), ::cuda::std::move(output), num_items, ::cuda::std::move(transform_op), stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Transforms one input sequence into one output sequence, by applying a transformation operation on corresponding @@ -120,7 +120,7 @@ struct DeviceTransform stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // This overload has additional parameters to specify temporary storage. Provided for compatibility with other CUB // APIs. template @@ -146,7 +146,7 @@ struct DeviceTransform ::cuda::std::move(transform_op), stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Overview @@ -189,7 +189,7 @@ struct DeviceTransform ::cuda::std::move(inputs), ::cuda::std::move(output), num_items, ::cuda::std::move(transform_op), stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_RUNTIME_FUNCTION static cudaError_t TransformStableArgumentAddresses( void* d_temp_storage, @@ -209,7 +209,7 @@ struct DeviceTransform return TransformStableArgumentAddresses( ::cuda::std::move(inputs), ::cuda::std::move(output), num_items, ::cuda::std::move(transform_op), stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //! @rst //! Transforms one input sequence into one output sequence, by applying a transformation operation on corresponding @@ -241,7 +241,7 @@ struct DeviceTransform stream); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template CUB_RUNTIME_FUNCTION static cudaError_t TransformStableArgumentAddresses( void* d_temp_storage, @@ -265,7 +265,7 @@ struct DeviceTransform ::cuda::std::move(transform_op), stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_adjacent_difference.cuh b/cub/cub/device/dispatch/dispatch_adjacent_difference.cuh index af41c7137c7..4eef4fb5b86 100644 --- a/cub/cub/device/dispatch/dispatch_adjacent_difference.cuh +++ b/cub/cub/device/dispatch/dispatch_adjacent_difference.cuh @@ -169,7 +169,7 @@ struct DispatchAdjacentDifference : public SelectedPolicy , stream(stream) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_DEPRECATED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchAdjacentDifference( void* d_temp_storage, @@ -190,7 +190,7 @@ struct DispatchAdjacentDifference : public SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /// Invocation template @@ -356,7 +356,7 @@ struct DispatchAdjacentDifference : public SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t Dispatch( void* d_temp_storage, @@ -372,7 +372,7 @@ struct DispatchAdjacentDifference : public SelectedPolicy return Dispatch(d_temp_storage, temp_storage_bytes, d_input, d_output, num_items, difference_op, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_histogram.cuh b/cub/cub/device/dispatch/dispatch_histogram.cuh index dab551559a4..15e0311fa2a 100644 --- a/cub/cub/device/dispatch/dispatch_histogram.cuh +++ b/cub/cub/device/dispatch/dispatch_histogram.cuh @@ -1036,7 +1036,7 @@ public: return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t DispatchRange( void* d_temp_storage, @@ -1067,7 +1067,7 @@ public: stream, is_byte_sample); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * Dispatch routine for HistogramRange, specialized for 8-bit sample types @@ -1202,7 +1202,7 @@ public: return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION static cudaError_t DispatchRange( void* d_temp_storage, @@ -1233,7 +1233,7 @@ public: stream, is_byte_sample); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * Dispatch routine for HistogramEven, specialized for sample types larger than 8-bit @@ -1420,7 +1420,7 @@ public: return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t DispatchEven( void* d_temp_storage, @@ -1453,7 +1453,7 @@ public: stream, is_byte_sample); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * Dispatch routine for HistogramEven, specialized for 8-bit sample types @@ -1592,7 +1592,7 @@ public: return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t DispatchEven( void* d_temp_storage, @@ -1625,7 +1625,7 @@ public: stream, is_byte_sample); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_reduce.cuh b/cub/cub/device/dispatch/dispatch_reduce.cuh index 23855d05951..c485e80e446 100644 --- a/cub/cub/device/dispatch/dispatch_reduce.cuh +++ b/cub/cub/device/dispatch/dispatch_reduce.cuh @@ -469,7 +469,7 @@ struct DispatchReduce : SelectedPolicy , launcher_factory(launcher_factory) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchReduce( void* d_temp_storage, @@ -494,7 +494,7 @@ struct DispatchReduce : SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //--------------------------------------------------------------------------- // Small-problem (single tile) invocation @@ -814,7 +814,7 @@ struct DispatchReduce : SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -831,7 +831,7 @@ struct DispatchReduce : SelectedPolicy return Dispatch(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items, reduction_op, init, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; /** @@ -1008,7 +1008,7 @@ struct DispatchSegmentedReduce : SelectedPolicy , ptx_version(ptx_version) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchSegmentedReduce( void* d_temp_storage, @@ -1037,7 +1037,7 @@ struct DispatchSegmentedReduce : SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //--------------------------------------------------------------------------- // Chained policy invocation @@ -1238,7 +1238,7 @@ struct DispatchSegmentedReduce : SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -1267,7 +1267,7 @@ struct DispatchSegmentedReduce : SelectedPolicy init, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_reduce_by_key.cuh b/cub/cub/device/dispatch/dispatch_reduce_by_key.cuh index 00d7280701a..482b9afe19f 100644 --- a/cub/cub/device/dispatch/dispatch_reduce_by_key.cuh +++ b/cub/cub/device/dispatch/dispatch_reduce_by_key.cuh @@ -550,7 +550,7 @@ struct DispatchReduceByKey return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -581,7 +581,7 @@ struct DispatchReduceByKey num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_rle.cuh b/cub/cub/device/dispatch/dispatch_rle.cuh index 2a6a0b3b641..bb99b20ab8a 100644 --- a/cub/cub/device/dispatch/dispatch_rle.cuh +++ b/cub/cub/device/dispatch/dispatch_rle.cuh @@ -543,7 +543,7 @@ struct DeviceRleDispatch return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -570,7 +570,7 @@ struct DeviceRleDispatch num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_scan.cuh b/cub/cub/device/dispatch/dispatch_scan.cuh index d1efaa01cd2..691fc2ece8c 100644 --- a/cub/cub/device/dispatch/dispatch_scan.cuh +++ b/cub/cub/device/dispatch/dispatch_scan.cuh @@ -330,7 +330,7 @@ struct DispatchScan : SelectedPolicy , ptx_version(ptx_version) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchScan( void* d_temp_storage, @@ -355,7 +355,7 @@ struct DispatchScan : SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED template CUB_RUNTIME_FUNCTION _CCCL_HOST _CCCL_FORCEINLINE cudaError_t Invoke(InitKernel init_kernel, ScanKernel scan_kernel) @@ -593,7 +593,7 @@ struct DispatchScan : SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -610,7 +610,7 @@ struct DispatchScan : SelectedPolicy return Dispatch(d_temp_storage, temp_storage_bytes, d_in, d_out, scan_op, init_value, num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_scan_by_key.cuh b/cub/cub/device/dispatch/dispatch_scan_by_key.cuh index aa04ce9f2ec..bf26c54e90e 100644 --- a/cub/cub/device/dispatch/dispatch_scan_by_key.cuh +++ b/cub/cub/device/dispatch/dispatch_scan_by_key.cuh @@ -342,7 +342,7 @@ struct DispatchScanByKey : SelectedPolicy , ptx_version(ptx_version) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchScanByKey( void* d_temp_storage, @@ -371,7 +371,7 @@ struct DispatchScanByKey : SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED template CUB_RUNTIME_FUNCTION _CCCL_HOST _CCCL_FORCEINLINE cudaError_t Invoke(InitKernel init_kernel, ScanKernel scan_kernel) @@ -622,7 +622,7 @@ struct DispatchScanByKey : SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -651,7 +651,7 @@ struct DispatchScanByKey : SelectedPolicy num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_segmented_sort.cuh b/cub/cub/device/dispatch/dispatch_segmented_sort.cuh index 80d8973c759..a98e1de494a 100644 --- a/cub/cub/device/dispatch/dispatch_segmented_sort.cuh +++ b/cub/cub/device/dispatch/dispatch_segmented_sort.cuh @@ -1131,7 +1131,7 @@ struct DispatchSegmentedSort : SelectedPolicy , stream(stream) {} -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE DispatchSegmentedSort( void* d_temp_storage, @@ -1158,7 +1158,7 @@ struct DispatchSegmentedSort : SelectedPolicy { CUB_DETAIL_RUNTIME_DEBUG_SYNC_USAGE_LOG } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED template CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE cudaError_t Invoke() @@ -1440,7 +1440,7 @@ struct DispatchSegmentedSort : SelectedPolicy return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -1469,7 +1469,7 @@ struct DispatchSegmentedSort : SelectedPolicy is_overwrite_okay, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE int GetNumPasses(int radix_bits) diff --git a/cub/cub/device/dispatch/dispatch_select_if.cuh b/cub/cub/device/dispatch/dispatch_select_if.cuh index 807ba62e4b3..7fbf9ccda4f 100644 --- a/cub/cub/device/dispatch/dispatch_select_if.cuh +++ b/cub/cub/device/dispatch/dispatch_select_if.cuh @@ -845,7 +845,7 @@ struct DispatchSelectIf : SelectedPolicy return CubDebug(MaxPolicyT::Invoke(ptx_version, dispatch)); } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document CUB_DETAIL_RUNTIME_DEBUG_SYNC_IS_NOT_SUPPORTED CUB_RUNTIME_FUNCTION _CCCL_FORCEINLINE static cudaError_t Dispatch( void* d_temp_storage, @@ -874,7 +874,7 @@ struct DispatchSelectIf : SelectedPolicy num_items, stream); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED }; CUB_NAMESPACE_END diff --git a/cub/cub/device/dispatch/dispatch_spmv_orig.cuh b/cub/cub/device/dispatch/dispatch_spmv_orig.cuh index a36a7f7890a..7d3d3094a48 100644 --- a/cub/cub/device/dispatch/dispatch_spmv_orig.cuh +++ b/cub/cub/device/dispatch/dispatch_spmv_orig.cuh @@ -893,7 +893,7 @@ struct DispatchSpmv return error; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template grid_queue, OffsetT n grid_queue.FillAndResetDrain(num_items); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/thread/thread_load.cuh b/cub/cub/thread/thread_load.cuh index 14577a56c92..6679f04b1e8 100644 --- a/cub/cub/thread/thread_load.cuh +++ b/cub/cub/thread/thread_load.cuh @@ -110,7 +110,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE cub::detail::value_t Thread //@} end member group -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Helper structure for templated load iteration (inductive case) /// \deprecated [Since 2.6.0] Use UnrolledThreadLoad() or UnrolledCopy() instead. @@ -378,6 +378,6 @@ _CCCL_DEVICE _CCCL_FORCEINLINE cub::detail::value_t Thread return ThreadLoad(itr, Int2Type(), Int2Type<::cuda::std::is_pointer::value>()); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/thread/thread_operators.cuh b/cub/cub/thread/thread_operators.cuh index 2de65083843..45d2446188f 100644 --- a/cub/cub/thread/thread_operators.cuh +++ b/cub/cub/thread/thread_operators.cuh @@ -396,7 +396,7 @@ CUB_DEPRECATED _CCCL_HOST_DEVICE BinaryFlip MakeBinaryFlip(BinaryOpT } _CCCL_SUPPRESS_DEPRECATED_POP -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace internal { @@ -720,6 +720,6 @@ using simd_type_t = typename CubOperatorToSimdOperator::simd_type; } // namespace internal -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/thread/thread_reduce.cuh b/cub/cub/thread/thread_reduce.cuh index 2a5b6566a26..d4b4a89fdfd 100644 --- a/cub/cub/thread/thread_reduce.cuh +++ b/cub/cub/thread/thread_reduce.cuh @@ -145,11 +145,11 @@ CUB_NAMESPACE_BEGIN //! template ()[0])>, #else typename ValueT = random_access_value_t, -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED typename AccumT = ::cuda::std::__accumulator_t> _CCCL_NODISCARD _CCCL_DEVICE _CCCL_FORCEINLINE AccumT ThreadReduce(const Input& input, ReductionOp reduction_op); // forward declaration @@ -158,7 +158,7 @@ _CCCL_NODISCARD _CCCL_DEVICE _CCCL_FORCEINLINE AccumT ThreadReduce(const Input& * Internal Reduction Implementations **********************************************************************************************************************/ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -697,6 +697,6 @@ _CCCL_NODISCARD _CCCL_DEVICE _CCCL_FORCEINLINE T ThreadReduce(const T*, Reductio } // namespace internal -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/thread/thread_store.cuh b/cub/cub/thread/thread_store.cuh index d0927a0d28d..a895884a60d 100644 --- a/cub/cub/thread/thread_store.cuh +++ b/cub/cub/thread/thread_store.cuh @@ -114,7 +114,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE void ThreadStore(OutputIteratorT itr, T val); //@} end member group -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Helper structure for templated store iteration (inductive case) template @@ -353,6 +353,6 @@ _CCCL_DEVICE _CCCL_FORCEINLINE void ThreadStore(OutputIteratorT itr, T val) ThreadStore(itr, val, Int2Type(), Int2Type::value>()); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/util_allocator.cuh b/cub/cub/util_allocator.cuh index d9559b874f3..39e59bdf4de 100644 --- a/cub/cub/util_allocator.cuh +++ b/cub/cub/util_allocator.cuh @@ -110,7 +110,7 @@ struct CachingDeviceAllocator /// Invalid size static constexpr size_t INVALID_SIZE = (size_t) -1; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Invalid device ordinal static constexpr int INVALID_DEVICE_ORDINAL = -1; @@ -299,7 +299,7 @@ struct CachingDeviceAllocator /// Set of live device allocations currently in use BusyBlocks live_blocks; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED //--------------------------------------------------------------------- // Methods diff --git a/cub/cub/util_arch.cuh b/cub/cub/util_arch.cuh index 5f8780620fa..1d6d7289b78 100644 --- a/cub/cub/util_arch.cuh +++ b/cub/cub/util_arch.cuh @@ -52,7 +52,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // \deprecated [Since 2.1.0] # define CUB_USE_COOPERATIVE_GROUPS diff --git a/cub/cub/util_cpp_dialect.cuh b/cub/cub/util_cpp_dialect.cuh index 006a070a7e9..6f54239bf84 100644 --- a/cub/cub/util_cpp_dialect.cuh +++ b/cub/cub/util_cpp_dialect.cuh @@ -42,7 +42,7 @@ #include // IWYU pragma: export -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // Deprecation warnings may be silenced by defining the following macros. These // may be combined. @@ -133,4 +133,4 @@ CUB_COMPILER_DEPRECATION_SOFT(C++ 17, C++ 14); # undef CUB_COMP_DEPR_IMPL0 # undef CUB_COMP_DEPR_IMPL1 -#endif // !DOXYGEN_SHOULD_SKIP_THIS +#endif // !_CCCL_DOXYGEN_INVOKED diff --git a/cub/cub/util_debug.cuh b/cub/cub/util_debug.cuh index edb75a64da3..0a08c9ae223 100644 --- a/cub/cub/util_debug.cuh +++ b/cub/cub/util_debug.cuh @@ -48,7 +48,7 @@ #include -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Only parse this during doxygen passes: +#ifdef _CCCL_DOXYGEN_INVOKED // Only parse this during doxygen passes: /** * @def CUB_DEBUG_LOG @@ -92,7 +92,7 @@ */ # define CUB_DEBUG_ALL -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED // `CUB_DETAIL_DEBUG_LEVEL_*`: Implementation details, internal use only: diff --git a/cub/cub/util_device.cuh b/cub/cub/util_device.cuh index e395b17f6d3..5b8c1f3f1f3 100644 --- a/cub/cub/util_device.cuh +++ b/cub/cub/util_device.cuh @@ -65,7 +65,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -90,7 +90,7 @@ template CUB_DETAIL_KERNEL_ATTRIBUTES void EmptyKernel() {} -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * \brief Returns the current device or -1 if an error occurred. @@ -105,13 +105,13 @@ CUB_RUNTIME_FUNCTION inline int CurrentDevice() return device; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document //! @brief RAII helper which saves the current device and switches to the specified device on construction and switches //! to the saved device on destruction. using SwitchDevice = ::cuda::__ensure_current_device; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * \brief Returns the number of CUDA devices available or -1 if an error @@ -171,7 +171,7 @@ CUB_RUNTIME_FUNCTION inline int DeviceCount() return result; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * \brief Per-device cache for a CUDA attribute value; the attribute is queried * and stored for each device upon construction. @@ -286,7 +286,7 @@ public: return entry.payload; } }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * \brief Retrieves the PTX version that will be used on the current device (major * 100 + minor * 10). diff --git a/cub/cub/util_macro.cuh b/cub/cub/util_macro.cuh index f98751b2ddf..b3ab7e73629 100644 --- a/cub/cub/util_macro.cuh +++ b/cub/cub/util_macro.cuh @@ -49,7 +49,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document # define CUB_PREVENT_MACRO_SUBSTITUTION template constexpr _CCCL_HOST_DEVICE auto min CUB_PREVENT_MACRO_SUBSTITUTION(T&& t, U&& u) diff --git a/cub/cub/util_ptx.cuh b/cub/cub/util_ptx.cuh index 3fc73b90304..aa522d9576e 100644 --- a/cub/cub/util_ptx.cuh +++ b/cub/cub/util_ptx.cuh @@ -97,7 +97,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE unsigned int SHL_ADD(unsigned int x, unsigned int return ret; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * Bitfield-extract. @@ -135,7 +135,7 @@ BFE(UnsignedBits source, unsigned int bit_start, unsigned int num_bits, Int2Type } # endif -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * \brief Bitfield-extract. Extracts \p num_bits from \p source starting at bit-offset \p bit_start. The input \p @@ -199,7 +199,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE int PRMT(unsigned int a, unsigned int b, unsigned return ret; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * Sync-threads barrier. @@ -329,7 +329,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE float FFMA_RZ(float a, float b, float c) return d; } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * \brief Terminates the calling thread @@ -689,7 +689,7 @@ _CCCL_DEVICE _CCCL_FORCEINLINE T ShuffleIndex(T input, int src_lane, unsigned in return output; } -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -751,7 +751,7 @@ struct warp_matcher_t }; } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * Compute a 32b mask of threads having the same least-significant diff --git a/cub/cub/util_temporary_storage.cuh b/cub/cub/util_temporary_storage.cuh index ee456083c3e..61c00f969f4 100644 --- a/cub/cub/util_temporary_storage.cuh +++ b/cub/cub/util_temporary_storage.cuh @@ -48,7 +48,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * @brief Alias temporaries to externally-allocated device storage (or simply return the amount of storage needed). @@ -112,6 +112,6 @@ _CCCL_HOST_DEVICE _CCCL_FORCEINLINE cudaError_t AliasTemporaries( return cudaSuccess; } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/util_type.cuh b/cub/cub/util_type.cuh index 42ffef0f6b0..f062ebc4ae9 100644 --- a/cub/cub/util_type.cuh +++ b/cub/cub/util_type.cuh @@ -85,7 +85,7 @@ CUB_NAMESPACE_BEGIN * Conditional types ******************************************************************************/ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { //! Alias to the given iterator's value_type. @@ -142,7 +142,7 @@ struct Log2 }; // Inductive case }; -# ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +# ifndef _CCCL_DOXYGEN_INVOKED // Do not document template struct Log2 @@ -155,7 +155,7 @@ struct Log2 }; }; -# endif // DOXYGEN_SHOULD_SKIP_THIS +# endif // _CCCL_DOXYGEN_INVOKED /** * \brief Statically determine if N is a power-of-two @@ -169,13 +169,13 @@ struct PowerOfTwo }; }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /****************************************************************************** * Marker types ******************************************************************************/ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /** * \brief A simple "null" marker type @@ -1156,6 +1156,6 @@ template struct Traits : NumericTraits::type> {}; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/util_vsmem.cuh b/cub/cub/util_vsmem.cuh index d2e5541c09c..f5926ce11e5 100644 --- a/cub/cub/util_vsmem.cuh +++ b/cub/cub/util_vsmem.cuh @@ -54,7 +54,7 @@ CUB_NAMESPACE_BEGIN -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace detail { @@ -248,6 +248,6 @@ using vsmem_helper_default_fallback_policy_t = } // namespace detail -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cub/cub/warp/warp_reduce.cuh b/cub/cub/warp/warp_reduce.cuh index 1d647a06c86..00440c18bdf 100644 --- a/cub/cub/warp/warp_reduce.cuh +++ b/cub/cub/warp/warp_reduce.cuh @@ -170,14 +170,14 @@ private: }; public: -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document /// Internal specialization. /// Use SHFL-based reduction if LOGICAL_WARP_THREADS is a power-of-two using InternalWarpReduce = ::cuda::std::_If, WarpReduceSmem>; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: /// Shared memory storage layout type for WarpReduce @@ -662,7 +662,7 @@ public: //! @} end member group }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template class WarpReduce { @@ -740,6 +740,6 @@ public: return input; } }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED CUB_NAMESPACE_END diff --git a/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh b/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh index 4bcd93d259f..5bfd60da9d3 100644 --- a/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh +++ b/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh @@ -273,12 +273,12 @@ public: __stream_ = __new_stream; } -# ifndef DOXYGEN_SHOULD_SKIP_THIS // friend functions are currently broken +# ifndef _CCCL_DOXYGEN_INVOKED // friend functions are currently broken //! @brief Forwards the passed properties _CCCL_TEMPLATE(class _Property) _CCCL_REQUIRES((!property_with_value<_Property>) _CCCL_AND _CUDA_VSTD::__is_included_in_v<_Property, _Properties...>) _CCCL_HIDE_FROM_ABI friend constexpr void get_property(const uninitialized_async_buffer&, _Property) noexcept {} -# endif // DOXYGEN_SHOULD_SKIP_THIS +# endif // _CCCL_DOXYGEN_INVOKED //! @brief Internal method to grow the allocation to a new size \p __count. //! @param __count The new size of the allocation. diff --git a/cudax/include/cuda/experimental/__container/uninitialized_buffer.cuh b/cudax/include/cuda/experimental/__container/uninitialized_buffer.cuh index d480ded4588..38c968d25c8 100644 --- a/cudax/include/cuda/experimental/__container/uninitialized_buffer.cuh +++ b/cudax/include/cuda/experimental/__container/uninitialized_buffer.cuh @@ -238,12 +238,12 @@ public: return __mr_; } -# ifndef DOXYGEN_SHOULD_SKIP_THIS // friend functions are currently broken +# ifndef _CCCL_DOXYGEN_INVOKED // friend functions are currently broken //! @brief Forwards the passed Properties _CCCL_TEMPLATE(class _Property) _CCCL_REQUIRES((!property_with_value<_Property>) _CCCL_AND _CUDA_VSTD::__is_included_in_v<_Property, _Properties...>) _CCCL_HIDE_FROM_ABI friend constexpr void get_property(const uninitialized_buffer&, _Property) noexcept {} -# endif // DOXYGEN_SHOULD_SKIP_THIS +# endif // _CCCL_DOXYGEN_INVOKED //! @brief Internal method to grow the allocation to a new size \p __count. //! @param __count The new size of the allocation. diff --git a/cudax/include/cuda/experimental/__device/device.cuh b/cudax/include/cuda/experimental/__device/device.cuh index 52c109bff6a..3e19bafb4e7 100644 --- a/cudax/include/cuda/experimental/__device/device.cuh +++ b/cudax/include/cuda/experimental/__device/device.cuh @@ -68,7 +68,7 @@ public: template <::cudaDeviceAttr _Attr> using attr_result_t = typename detail::__dev_attr<_Attr>::type; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document # if defined(_CCCL_COMPILER_MSVC) // When __EDG__ is defined, std::construct_at will not permit constructing // a device object from an __emplace_device object. This is a workaround. diff --git a/cudax/include/cuda/experimental/__memory_resource/device_memory_resource.cuh b/cudax/include/cuda/experimental/__memory_resource/device_memory_resource.cuh index bae301feb0b..7d54dd4f750 100644 --- a/cudax/include/cuda/experimental/__memory_resource/device_memory_resource.cuh +++ b/cudax/include/cuda/experimental/__memory_resource/device_memory_resource.cuh @@ -405,11 +405,11 @@ public: return __pool_; } -# ifndef DOXYGEN_SHOULD_SKIP_THIS // Doxygen cannot handle the friend function +# ifndef _CCCL_DOXYGEN_INVOKED // Doxygen cannot handle the friend function //! @brief Enables the \c device_accessible property for \c device_memory_resource. //! @relates device_memory_resource friend constexpr void get_property(device_memory_resource const&, _CUDA_VMR::device_accessible) noexcept {} -# endif // DOXYGEN_SHOULD_SKIP_THIS +# endif // _CCCL_DOXYGEN_INVOKED }; static_assert(_CUDA_VMR::resource_with, ""); diff --git a/cudax/include/cuda/experimental/__stf/internal/data_interface.cuh b/cudax/include/cuda/experimental/__stf/internal/data_interface.cuh index 0d2026fdbbe..7fe81211569 100644 --- a/cudax/include/cuda/experimental/__stf/internal/data_interface.cuh +++ b/cudax/include/cuda/experimental/__stf/internal/data_interface.cuh @@ -213,7 +213,7 @@ public: */ virtual size_t data_hash(instance_id_t instance_id) const = 0; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen fails to parse this +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen fails to parse this /** * @brief Returns the size of the data represented by this logical data. * @@ -221,7 +221,7 @@ public: * purposes, or for the scheduling strategies. */ virtual size_t data_footprint() const = 0; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Get the part of the data interface that is common to all data instances. diff --git a/cudax/include/cuda/experimental/__stf/internal/execution_policy.cuh b/cudax/include/cuda/experimental/__stf/internal/execution_policy.cuh index d2cc954bfa9..e79dca54141 100644 --- a/cudax/include/cuda/experimental/__stf/internal/execution_policy.cuh +++ b/cudax/include/cuda/experimental/__stf/internal/execution_policy.cuh @@ -301,7 +301,7 @@ public: * * @tparam level The level in the hierarchy to check for the `sync` property. Level starts from 0 (top-level). */ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen fails to parse this +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen fails to parse this template static inline constexpr bool is_synchronizable = [] { if constexpr (level > 0) @@ -395,7 +395,7 @@ private: mem mem_bytes = mem(0); }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen fails to parse this +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen fails to parse this /** * @brief Creates and returns a `thread_hierarchy_spec` object with no synchronization and dynamic width. * @@ -480,7 +480,7 @@ constexpr auto con(const P&... p) return R(p...); } /// @} -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED #ifdef UNITTESTED_FILE diff --git a/cudax/include/cuda/experimental/__stf/internal/reduction_base.cuh b/cudax/include/cuda/experimental/__stf/internal/reduction_base.cuh index 0b6a0cd7c78..1d8d00d6670 100644 --- a/cudax/include/cuda/experimental/__stf/internal/reduction_base.cuh +++ b/cudax/include/cuda/experimental/__stf/internal/reduction_base.cuh @@ -42,7 +42,7 @@ public: reduction_operator_base& operator=(const reduction_operator_base&) = delete; reduction_operator_base(const reduction_operator_base&) = delete; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen fails here +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen fails here // Reduction operator (inout, in) virtual void op_untyped( @@ -62,7 +62,7 @@ public: const exec_place& e, event_list& prereq_in) = 0; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED private: // not used for now ... diff --git a/cudax/include/cuda/experimental/__stf/places/exec/host/callback_queues.cuh b/cudax/include/cuda/experimental/__stf/places/exec/host/callback_queues.cuh index 0b011cce0f5..2d3036ec143 100644 --- a/cudax/include/cuda/experimental/__stf/places/exec/host/callback_queues.cuh +++ b/cudax/include/cuda/experimental/__stf/places/exec/host/callback_queues.cuh @@ -30,7 +30,7 @@ #include #include -#ifndef DOXYGEN_SHOULD_SKIP_THIS // do not document +#ifndef _CCCL_DOXYGEN_INVOKED // do not document # if !defined(_CCCL_COMPILER_MSVC) # define STATEFUL_CALLBACKS @@ -603,4 +603,4 @@ inline bool cudaCallbackQueueProgress(callback_queue* q, bool flag) } // end namespace cuda::experimental::stf # endif // !_CCCL_COMPILER_MSVC -#endif // DOXYGEN_SHOULD_SKIP_THIS do not document +#endif // _CCCL_DOXYGEN_INVOKED do not document diff --git a/cudax/include/cuda/experimental/__stf/places/inner_shape.cuh b/cudax/include/cuda/experimental/__stf/places/inner_shape.cuh index 04b2badf7f2..383f43961be 100644 --- a/cudax/include/cuda/experimental/__stf/places/inner_shape.cuh +++ b/cudax/include/cuda/experimental/__stf/places/inner_shape.cuh @@ -31,7 +31,7 @@ namespace cuda::experimental::stf { -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen fails to parse this +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen fails to parse this /** * @brief Applying "inner" on a mdspan shape returns an explicit shape which extents @@ -89,7 +89,7 @@ _CCCL_HOST_DEVICE box inner(const box& s) return box(inner_extents); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED #ifdef UNITTESTED_FILE UNITTEST("inner explicit shape (explicit bounds)") diff --git a/cudax/include/cuda/experimental/__stf/stream/reduction.cuh b/cudax/include/cuda/experimental/__stf/stream/reduction.cuh index 4493672c70b..deea02bbd9c 100644 --- a/cudax/include/cuda/experimental/__stf/stream/reduction.cuh +++ b/cudax/include/cuda/experimental/__stf/stream/reduction.cuh @@ -65,7 +65,7 @@ public: const exec_place& e, cudaStream_t s) = 0; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen has issues with this code +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen has issues with this code void op_untyped( logical_data_untyped& d, const data_place& inout_memory_node, @@ -110,7 +110,7 @@ public: prereqs = async_op.end(d.get_ctx()); } -#endif // DOXYGEN_SHOULD_SKIP_THIS // doxygen has issues with this code +#endif // _CCCL_DOXYGEN_INVOKED // doxygen has issues with this code }; /** diff --git a/cudax/include/cuda/experimental/__stf/stream/stream_task.cuh b/cudax/include/cuda/experimental/__stf/stream/stream_task.cuh index 3c51f7304bb..48a28aa6648 100644 --- a/cudax/include/cuda/experimental/__stf/stream/stream_task.cuh +++ b/cudax/include/cuda/experimental/__stf/stream/stream_task.cuh @@ -617,7 +617,7 @@ private: template class deferred_stream_task; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // doxygen has issues with this code +#ifndef _CCCL_DOXYGEN_INVOKED // doxygen has issues with this code /* * Base of all deferred tasks. Stores the needed information for typed deferred tasks to run (see below). */ @@ -877,6 +877,6 @@ public: }; } }; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED } // namespace cuda::experimental::stf diff --git a/cudax/include/cuda/experimental/__stf/utility/core.cuh b/cudax/include/cuda/experimental/__stf/utility/core.cuh index e0eb417aad7..23b0ff5560f 100644 --- a/cudax/include/cuda/experimental/__stf/utility/core.cuh +++ b/cudax/include/cuda/experimental/__stf/utility/core.cuh @@ -79,7 +79,7 @@ inline int setenv(const char* name, const char* value, int overwrite) } #endif -#ifndef DOXYGEN_SHOULD_SKIP_THIS // FIXME Doxygen is lost with decltype(auto) +#ifndef _CCCL_DOXYGEN_INVOKED // FIXME Doxygen is lost with decltype(auto) /** * @brief Custom move function that performs checks on the argument type. * @@ -97,7 +97,7 @@ _CCCL_HOST_DEVICE constexpr decltype(auto) mv(T&& obj) static_assert(!::std::is_const_v<::std::remove_reference_t>, "Misleading move from const lvalue."); return ::std::move(obj); } -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /** * @brief Creates a `std::shared_ptr` managing a copy of the given object. @@ -609,7 +609,7 @@ private: [[no_unique_address]] state_t payload = state_t(); }; -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // Operator implementations # define _3197bc91feaf98030b2cc0b441d7b0ea(op) \ template \ @@ -691,6 +691,6 @@ _3197bc91feaf98030b2cc0b441d7b0ea(>=); # undef _3197bc91feaf98030b2cc0b441d7b0ea -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED } // namespace cuda::experimental::stf diff --git a/cudax/include/cuda/experimental/__stf/utility/unittest.cuh b/cudax/include/cuda/experimental/__stf/utility/unittest.cuh index 3cc470df80e..dd42fbdd9bd 100644 --- a/cudax/include/cuda/experimental/__stf/utility/unittest.cuh +++ b/cudax/include/cuda/experimental/__stf/utility/unittest.cuh @@ -31,7 +31,7 @@ #include -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document // One level of macro indirection is required in order to resolve __COUNTER__, // and get varname1 instead of varname__COUNTER__. # define _55f56f4e3b45c8cf3fa50b28fed72e2a(a, b) _a56ec7069122ad2e0888a508ecdc4639(a, b) @@ -705,7 +705,7 @@ UNITTEST("cuda::std::source_location") test_func(); }; -#else // DOXYGEN_SHOULD_SKIP_THIS Do not document +#else // _CCCL_DOXYGEN_INVOKED Do not document // Ensure these are ignored by Doxygen # define UNITTEST(name, ...) -#endif // DOXYGEN_SHOULD_SKIP_THIS Do not document +#endif // _CCCL_DOXYGEN_INVOKED Do not document diff --git a/cudax/include/cuda/experimental/__utility/ensure_current_device.cuh b/cudax/include/cuda/experimental/__utility/ensure_current_device.cuh index 6c37d4f6996..c644dd19a1c 100644 --- a/cudax/include/cuda/experimental/__utility/ensure_current_device.cuh +++ b/cudax/include/cuda/experimental/__utility/ensure_current_device.cuh @@ -27,7 +27,7 @@ #include #include -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document namespace cuda::experimental { @@ -101,5 +101,5 @@ struct [[maybe_unused]] __ensure_current_device } }; } // namespace cuda::experimental -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED #endif // _CUDAX__UTILITY_ENSURE_CURRENT_DEVICE diff --git a/docs/repo.toml b/docs/repo.toml index 9a684c4d5f4..f7c426f13db 100644 --- a/docs/repo.toml +++ b/docs/repo.toml @@ -159,11 +159,10 @@ doxygen_predefined = [ "_CCCL_DIAG_SUPPRESS_ICC(x)=", "_CCCL_DIAG_SUPPRESS_MSVC(x)=", "_CCCL_DIAG_SUPPRESS_NVHPC(x)=", + "_CCCL_DOXYGEN_INVOKED", "_CCCL_REQUIRES(x)= ::cuda::std::enable_if_t = 0>", "_CCCL_TEMPLATE(x)=template x _CCCL_EAT_REST", - "DOXYGEN_SHOULD_SKIP_THIS", - "DOXYGEN_ACTIVE", "__device__", "__host__", "__forceinline__", @@ -275,8 +274,7 @@ doxygen_predefined = [ "CUDASTF_HOST=", "CUDASTF_DEVICE=", "CUDASTF_HOST_DEVICE=", - "DOXYGEN_SHOULD_SKIP_THIS", - "DOXYGEN_ACTIVE", + "_CCCL_DOXYGEN_INVOKED", "_LIBCUDACXX_DEPRECATED_IN_CXX11", "THRUST_DISABLE_NAMESPACE_MAGIC", "THRUST_IGNORE_NAMESPACE_MAGIC_ERROR", @@ -445,8 +443,7 @@ doxygen_predefined = [ "_CUDAX_TRIVIAL_DEVICE_API", "_CUDAX_PUBLIC_API", "LIBCUDACXX_ENABLE_EXPERIMENTAL_MEMORY_RESOURCE=", - "DOXYGEN_SHOULD_SKIP_THIS", - "DOXYGEN_ACTIVE", + "_CCCL_DOXYGEN_INVOKED", ] # make sure to use ./fetch_imgs.sh diff --git a/libcudacxx/include/cuda/std/__type_traits/type_list.h b/libcudacxx/include/cuda/std/__type_traits/type_list.h index bef58f29966..4bd928b0013 100644 --- a/libcudacxx/include/cuda/std/__type_traits/type_list.h +++ b/libcudacxx/include/cuda/std/__type_traits/type_list.h @@ -42,7 +42,7 @@ _LIBCUDACXX_BEGIN_NAMESPACE_STD -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template struct __type_list; @@ -947,7 +947,7 @@ template using __type_iota = decltype(__detail::__type_iota_fn<_Ty, _Start, _Stride>(static_cast*>(nullptr))); -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED _LIBCUDACXX_END_NAMESPACE_STD diff --git a/thrust/thrust/detail/type_deduction.h b/thrust/thrust/detail/type_deduction.h index 08f31630bb5..a1d41de9676 100644 --- a/thrust/thrust/detail/type_deduction.h +++ b/thrust/thrust/detail/type_deduction.h @@ -59,7 +59,7 @@ /// // Trailing return types seem to confuse Doxygen, and cause it to interpret // parts of the function's body as new function signatures. -#if defined(THRUST_DOXYGEN) +#if defined(_CCCL_DOXYGEN_INVOKED) # define THRUST_DECLTYPE_RETURNS(...) \ { \ return (__VA_ARGS__); \ @@ -81,7 +81,7 @@ /// // Trailing return types seem to confuse Doxygen, and cause it to interpret // parts of the function's body as new function signatures. -#if defined(THRUST_DOXYGEN) +#if defined(_CCCL_DOXYGEN_INVOKED) # define THRUST_DECLTYPE_RETURNS(...) \ { \ return (__VA_ARGS__); \ diff --git a/thrust/thrust/device_malloc_allocator.h b/thrust/thrust/device_malloc_allocator.h index e5d2e04fc19..c9de52a8404 100644 --- a/thrust/thrust/device_malloc_allocator.h +++ b/thrust/thrust/device_malloc_allocator.h @@ -40,12 +40,12 @@ THRUST_NAMESPACE_BEGIN // forward declarations to WAR circular #includes -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document +#ifndef _CCCL_DOXYGEN_INVOKED // Do not document template class device_ptr; template device_ptr device_malloc(const std::size_t n); -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! \addtogroup allocators Allocators * \ingroup memory_management diff --git a/thrust/thrust/device_ptr.h b/thrust/thrust/device_ptr.h index 5c5f55a3a83..058d12cb83f 100644 --- a/thrust/thrust/device_ptr.h +++ b/thrust/thrust/device_ptr.h @@ -154,14 +154,14 @@ class device_ptr return *this; } -#if THRUST_DOXYGEN +#ifdef _CCCL_DOXYGEN_INVOKED /*! \brief Return the raw pointer that this \c device_ptr points to. */ _CCCL_HOST_DEVICE T* get() const; #endif }; -#if THRUST_DOXYGEN +#ifdef _CCCL_DOXYGEN_INVOKED /*! Write the address that a \c device_ptr points to to an output stream. * * \param os The output stream. diff --git a/thrust/thrust/device_reference.h b/thrust/thrust/device_reference.h index 40a6790a5a1..545d5449bee 100644 --- a/thrust/thrust/device_reference.h +++ b/thrust/thrust/device_reference.h @@ -961,7 +961,7 @@ _CCCL_HOST_DEVICE void swap(device_reference& x, device_reference& y) // declare these methods for the purpose of Doxygenating them // they actually are defined for a base class -#if THRUST_DOXYGEN +#ifdef _CCCL_DOXYGEN_INVOKED /*! Writes to an output stream the value of a \p device_reference. * * \param os The output stream. diff --git a/thrust/thrust/memory.h b/thrust/thrust/memory.h index 6462545590b..290c99b7b2e 100644 --- a/thrust/thrust/memory.h +++ b/thrust/thrust/memory.h @@ -138,7 +138,7 @@ template _CCCL_HOST_DEVICE pointer malloc(const thrust::detail::execution_policy_base& system, std::size_t n); -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! This version of \p malloc allocates typed uninitialized storage associated with a given system. * diff --git a/thrust/thrust/optional.h b/thrust/thrust/optional.h index 6762271cb47..bb9bf1cfb4b 100644 --- a/thrust/thrust/optional.h +++ b/thrust/thrust/optional.h @@ -1976,7 +1976,7 @@ optional(T) -> optional; #endif // Doxygen chokes on the trailing return types used below. -#if !defined(THRUST_DOXYGEN) +#if !defined(_CCCL_DOXYGEN_INVOKED) /// \exclude namespace detail { @@ -2034,7 +2034,7 @@ _CCCL_HOST_DEVICE auto optional_map_impl(Opt&& opt, F&& f) -> optional` acts similarly /// to a `T*`, but provides more operations and shows intent more clearly. diff --git a/thrust/thrust/pair.h b/thrust/thrust/pair.h index e3c74677993..9f35a388bc7 100644 --- a/thrust/thrust/pair.h +++ b/thrust/thrust/pair.h @@ -49,12 +49,12 @@ THRUST_NAMESPACE_BEGIN * \tparam N This parameter selects the member of interest. * \tparam T A \c pair type of interest. */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using tuple_element = _CUDA_VSTD::tuple_element; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::tuple_element; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! This convenience metafunction is included for compatibility with * \p tuple. It returns \c 2, the number of elements of a \p pair, @@ -62,12 +62,12 @@ using _CUDA_VSTD::tuple_element; * * \tparam Pair A \c pair type of interest. */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using tuple_size = _CUDA_VSTD::tuple_size; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::tuple_size; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! \p pair is a generic data structure encapsulating a heterogeneous * pair of values. @@ -80,12 +80,12 @@ using _CUDA_VSTD::tuple_size; * requirements on the type of \p T2. T2's type is * provided by pair::second_type. */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using pair = _CUDA_VSTD::pair; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::pair; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED using _CUDA_VSTD::get; using _CUDA_VSTD::make_pair; diff --git a/thrust/thrust/random/linear_congruential_engine.h b/thrust/thrust/random/linear_congruential_engine.h index ce47c08b619..c289667749f 100644 --- a/thrust/thrust/random/linear_congruential_engine.h +++ b/thrust/thrust/random/linear_congruential_engine.h @@ -143,11 +143,11 @@ class linear_congruential_engine /*! The smallest value this \p linear_congruential_engine may potentially produce. */ -#ifndef DOXYGEN_SHOULD_SKIP_THIS // Doxygen breaks on the ternary :shrug: +#ifndef _CCCL_DOXYGEN_INVOKED // Doxygen breaks on the ternary :shrug: static const result_type min = c == 0u ? 1u : 0u; #else static const result_type min = 0u; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! The largest value this \p linear_congruential_engine may potentially produce. */ diff --git a/thrust/thrust/tuple.h b/thrust/thrust/tuple.h index 1f8ed8943e5..d0d13670f0c 100644 --- a/thrust/thrust/tuple.h +++ b/thrust/thrust/tuple.h @@ -94,12 +94,12 @@ _CCCL_HOST_DEVICE inline bool operator>(const null_type&, const null_type&) * \see pair * \see tuple */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using tuple_element = _CUDA_VSTD::tuple_element; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::tuple_element; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! This metafunction returns the number of elements * of a \p tuple type of interest. @@ -109,12 +109,12 @@ using _CUDA_VSTD::tuple_element; * \see pair * \see tuple */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using tuple_size = _CUDA_VSTD::tuple_size; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::tuple_size; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED /*! \brief \p tuple is a heterogeneous, fixed-size collection of values. * An instantiation of \p tuple with two arguments is similar to an @@ -153,12 +153,12 @@ using _CUDA_VSTD::tuple_size; * \see tuple_size * \see tie */ -#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen +#ifdef _CCCL_DOXYGEN_INVOKED // Provide a fake alias for doxygen template using tuple = _CUDA_VSTD::tuple; -#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +#else // ^^^ _CCCL_DOXYGEN_INVOKED ^^^ / vvv !_CCCL_DOXYGEN_INVOKED vvv using _CUDA_VSTD::tuple; -#endif // DOXYGEN_SHOULD_SKIP_THIS +#endif // _CCCL_DOXYGEN_INVOKED using _CUDA_VSTD::get; using _CUDA_VSTD::make_tuple;