OpenFPM_pdata  4.1.0
Project that contain the implementation of distributed structures
 
Loading...
Searching...
No Matches
Data Structures | Public Types | Static Public Member Functions
cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT > Struct Template Reference

< Signed integer type for global offsets More...

Detailed Description

template<typename KeysInputIteratorT, typename UniqueOutputIteratorT, typename ValuesInputIteratorT, typename AggregatesOutputIteratorT, typename NumRunsOutputIteratorT, typename EqualityOpT, typename ReductionOpT, typename OffsetT>
struct cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >

< Signed integer type for global offsets

Utility class for dispatching the appropriately-tuned kernels for DeviceReduceByKey

Definition at line 126 of file dispatch_reduce_by_key.cuh.

Data Structures

struct  KernelConfig
 
struct  Policy110
 SM11. More...
 
struct  Policy130
 SM13. More...
 
struct  Policy200
 SM20. More...
 
struct  Policy300
 SM30. More...
 
struct  Policy350
 SM35. More...
 
struct  PtxReduceByKeyPolicy
 

Public Types

enum  { INIT_KERNEL_THREADS = 128 , MAX_INPUT_BYTES = CUB_MAX(sizeof(KeyOutputT), sizeof(ValueOutputT)) , COMBINED_INPUT_BYTES = sizeof(KeyOutputT) + sizeof(ValueOutputT) }
 
typedef std::iterator_traits< KeysInputIteratorT >::value_type KeyInputT
 
typedef If<(Equals< typenamestd::iterator_traits< UniqueOutputIteratorT >::value_type, void >::VALUE), typenamestd::iterator_traits< KeysInputIteratorT >::value_type, typenamestd::iterator_traits< UniqueOutputIteratorT >::value_type >::Type KeyOutputT
 
typedef std::iterator_traits< ValuesInputIteratorT >::value_type ValueInputT
 
typedef If<(Equals< typenamestd::iterator_traits< AggregatesOutputIteratorT >::value_type, void >::VALUE), typenamestd::iterator_traits< ValuesInputIteratorT >::value_type, typenamestd::iterator_traits< AggregatesOutputIteratorT >::value_type >::Type ValueOutputT
 
typedef ReduceByKeyScanTileState< ValueOutputT, OffsetT > ScanTileStateT
 
typedef Policy110 PtxPolicy
 

Static Public Member Functions

template<typename KernelConfig >
CUB_RUNTIME_FUNCTION static __forceinline__ void InitConfigs (int ptx_version, KernelConfig &reduce_by_key_config)
 
template<typename ScanInitKernelT , typename ReduceByKeyKernelT >
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t Dispatch (void *d_temp_storage, size_t &temp_storage_bytes, KeysInputIteratorT d_keys_in, UniqueOutputIteratorT d_unique_out, ValuesInputIteratorT d_values_in, AggregatesOutputIteratorT d_aggregates_out, NumRunsOutputIteratorT d_num_runs_out, EqualityOpT equality_op, ReductionOpT reduction_op, OffsetT num_items, cudaStream_t stream, bool debug_synchronous, int, ScanInitKernelT init_kernel, ReduceByKeyKernelT reduce_by_key_kernel, KernelConfig reduce_by_key_config)
 < Function type of cub::DeviceReduceByKeyKernelT
 
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t Dispatch (void *d_temp_storage, size_t &temp_storage_bytes, KeysInputIteratorT d_keys_in, UniqueOutputIteratorT d_unique_out, ValuesInputIteratorT d_values_in, AggregatesOutputIteratorT d_aggregates_out, NumRunsOutputIteratorT d_num_runs_out, EqualityOpT equality_op, ReductionOpT reduction_op, OffsetT num_items, cudaStream_t stream, bool debug_synchronous)
 

Member Typedef Documentation

◆ KeyInputT

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef std::iterator_traits<KeysInputIteratorT>::value_type cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::KeyInputT

Definition at line 133 of file dispatch_reduce_by_key.cuh.

◆ KeyOutputT

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef If<(Equals<typenamestd::iterator_traits<UniqueOutputIteratorT>::value_type,void>::VALUE),typenamestd::iterator_traits<KeysInputIteratorT>::value_type,typenamestd::iterator_traits<UniqueOutputIteratorT>::value_type>::Type cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::KeyOutputT

Definition at line 138 of file dispatch_reduce_by_key.cuh.

◆ PtxPolicy

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef Policy110 cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::PtxPolicy

Definition at line 266 of file dispatch_reduce_by_key.cuh.

◆ ScanTileStateT

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef ReduceByKeyScanTileState<ValueOutputT, OffsetT> cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::ScanTileStateT

Definition at line 156 of file dispatch_reduce_by_key.cuh.

◆ ValueInputT

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef std::iterator_traits<ValuesInputIteratorT>::value_type cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::ValueInputT

Definition at line 141 of file dispatch_reduce_by_key.cuh.

◆ ValueOutputT

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
typedef If<(Equals<typenamestd::iterator_traits<AggregatesOutputIteratorT>::value_type,void>::VALUE),typenamestd::iterator_traits<ValuesInputIteratorT>::value_type,typenamestd::iterator_traits<AggregatesOutputIteratorT>::value_type>::Type cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::ValueOutputT

Definition at line 146 of file dispatch_reduce_by_key.cuh.

Member Enumeration Documentation

◆ anonymous enum

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
anonymous enum

Definition at line 148 of file dispatch_reduce_by_key.cuh.

Member Function Documentation

◆ Dispatch() [1/2]

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::Dispatch ( void *  d_temp_storage,
size_t &  temp_storage_bytes,
KeysInputIteratorT  d_keys_in,
UniqueOutputIteratorT  d_unique_out,
ValuesInputIteratorT  d_values_in,
AggregatesOutputIteratorT  d_aggregates_out,
NumRunsOutputIteratorT  d_num_runs_out,
EqualityOpT  equality_op,
ReductionOpT  reduction_op,
OffsetT  num_items,
cudaStream_t  stream,
bool  debug_synchronous 
)
inlinestatic

Internal dispatch routine

Parameters
[in]d_temp_storageDevice-accessible allocation of temporary storage. When NULL, the required allocation size is written to temp_storage_bytes and no work is done.
[in,out]temp_storage_bytesReference to size in bytes of d_temp_storage allocation
[in]d_keys_inPointer to the input sequence of keys
[out]d_unique_outPointer to the output sequence of unique keys (one key per run)
[in]d_values_inPointer to the input sequence of corresponding values
[out]d_aggregates_outPointer to the output sequence of value aggregates (one aggregate per run)
[out]d_num_runs_outPointer to total number of runs encountered (i.e., the length of d_unique_out)
[in]equality_opKeyT equality operator
[in]reduction_opValueT reduction operator
[in]num_itemsTotal number of items to select from
[in]streamCUDA stream to launch kernels within. Default is stream0.
[in]debug_synchronousWhether or not to synchronize the stream after every kernel launch to check for errors. Also causes launch configurations to be printed to the console. Default is false.

Definition at line 497 of file dispatch_reduce_by_key.cuh.

◆ Dispatch() [2/2]

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
template<typename ScanInitKernelT , typename ReduceByKeyKernelT >
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::Dispatch ( void *  d_temp_storage,
size_t &  temp_storage_bytes,
KeysInputIteratorT  d_keys_in,
UniqueOutputIteratorT  d_unique_out,
ValuesInputIteratorT  d_values_in,
AggregatesOutputIteratorT  d_aggregates_out,
NumRunsOutputIteratorT  d_num_runs_out,
EqualityOpT  equality_op,
ReductionOpT  reduction_op,
OffsetT  num_items,
cudaStream_t  stream,
bool  debug_synchronous,
int  ,
ScanInitKernelT  init_kernel,
ReduceByKeyKernelT  reduce_by_key_kernel,
KernelConfig  reduce_by_key_config 
)
inlinestatic

< Function type of cub::DeviceReduceByKeyKernelT

Internal dispatch routine for computing a device-wide reduce-by-key using the specified kernel functions.

Parameters
[in]d_temp_storageDevice-accessible allocation of temporary storage. When NULL, the required allocation size is written to temp_storage_bytes and no work is done.
[in,out]temp_storage_bytesReference to size in bytes of d_temp_storage allocation
[in]d_keys_inPointer to the input sequence of keys
[out]d_unique_outPointer to the output sequence of unique keys (one key per run)
[in]d_values_inPointer to the input sequence of corresponding values
[out]d_aggregates_outPointer to the output sequence of value aggregates (one aggregate per run)
[out]d_num_runs_outPointer to total number of runs encountered (i.e., the length of d_unique_out)
[in]equality_opKeyT equality operator
[in]reduction_opValueT reduction operator
[in]num_itemsTotal number of items to select from
[in]streamCUDA stream to launch kernels within. Default is stream0.
[in]debug_synchronousWhether or not to synchronize the stream after every kernel launch to check for errors. Also causes launch configurations to be printed to the console. Default is false.
[in]init_kernelKernel function pointer to parameterization of cub::DeviceScanInitKernel
[in]reduce_by_key_kernelKernel function pointer to parameterization of cub::DeviceReduceByKeyKernel
[in]reduce_by_key_configDispatch parameters that match the policy that reduce_by_key_kernel was compiled for

Definition at line 353 of file dispatch_reduce_by_key.cuh.

◆ InitConfigs()

template<typename KeysInputIteratorT , typename UniqueOutputIteratorT , typename ValuesInputIteratorT , typename AggregatesOutputIteratorT , typename NumRunsOutputIteratorT , typename EqualityOpT , typename ReductionOpT , typename OffsetT >
template<typename KernelConfig >
CUB_RUNTIME_FUNCTION static __forceinline__ void cub::DispatchReduceByKey< KeysInputIteratorT, UniqueOutputIteratorT, ValuesInputIteratorT, AggregatesOutputIteratorT, NumRunsOutputIteratorT, EqualityOpT, ReductionOpT, OffsetT >::InitConfigs ( int  ptx_version,
KernelConfig &  reduce_by_key_config 
)
inlinestatic

Initialize kernel dispatch configurations with the policies corresponding to the PTX assembly we will use

Definition at line 283 of file dispatch_reduce_by_key.cuh.

The documentation for this struct was generated from the following file: