OpenFPM_pdata  4.1.0
Project that contain the implementation of distributed structures
 
Loading...
Searching...
No Matches
Data Structures | Public Types | Public Member Functions | Static Public Member Functions | Data Fields
cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT > Struct Template Reference

< Signed integer type for global offsets More...

Detailed Description

template<bool IS_DESCENDING, typename KeyT, typename ValueT, typename OffsetT>
struct cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >

< Signed integer type for global offsets

Utility class for dispatching the appropriately-tuned kernels for device-wide radix sort

Definition at line 807 of file dispatch_radix_sort.cuh.

+ Inheritance diagram for cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >:

Data Structures

struct  PassConfig
 Pass configuration structure. More...
 

Public Types

enum  { KEYS_ONLY = (Equals<ValueT, NullType>::VALUE) }
 
- Public Types inherited from cub::DeviceRadixSortPolicy< KeyT, ValueT, OffsetT >
enum  { KEYS_ONLY = (Equals<ValueT, NullType>::VALUE) }
 
typedef If<(sizeof(ValueT)>4)&&(sizeof(KeyT)< sizeof(ValueT)), ValueT, KeyT >::Type DominantT
 
typedef Policy700 MaxPolicy
 MaxPolicy.
 

Public Member Functions

CUB_RUNTIME_FUNCTION __forceinline__ DispatchRadixSort (void *d_temp_storage, size_t &temp_storage_bytes, DoubleBuffer< KeyT > &d_keys, DoubleBuffer< ValueT > &d_values, OffsetT num_items, int begin_bit, int end_bit, bool is_overwrite_okay, cudaStream_t stream, bool debug_synchronous, int ptx_version)
 Constructor.
 
template<typename ActivePolicyT , typename SingleTileKernelT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t InvokeSingleTile (SingleTileKernelT single_tile_kernel)
 Invoke a single block to sort in-core.
 
template<typename PassConfigT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t InvokePass (const KeyT *d_keys_in, KeyT *d_keys_out, const ValueT *d_values_in, ValueT *d_values_out, OffsetT *d_spine, int spine_length, int &current_bit, PassConfigT &pass_config)
 
template<typename ActivePolicyT , typename UpsweepKernelT , typename ScanKernelT , typename DownsweepKernelT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t InvokePasses (UpsweepKernelT upsweep_kernel, UpsweepKernelT alt_upsweep_kernel, ScanKernelT scan_kernel, DownsweepKernelT downsweep_kernel, DownsweepKernelT alt_downsweep_kernel)
 Invocation (run multiple digit passes)
 
template<typename ActivePolicyT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t Invoke ()
 Invocation.
 

Static Public Member Functions

CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t Dispatch (void *d_temp_storage, size_t &temp_storage_bytes, DoubleBuffer< KeyT > &d_keys, DoubleBuffer< ValueT > &d_values, OffsetT num_items, int begin_bit, int end_bit, bool is_overwrite_okay, cudaStream_t stream, bool debug_synchronous)
 

Data Fields

void * d_temp_storage
 [in] Device-accessible allocation of temporary storage. When NULL, the required allocation size is written to temp_storage_bytes and no work is done.
 
size_t & temp_storage_bytes
 [in,out] Reference to size in bytes of d_temp_storage allocation
 
DoubleBuffer< KeyT > & d_keys
 [in,out] Double-buffer whose current buffer contains the unsorted input keys and, upon return, is updated to point to the sorted output keys
 
DoubleBuffer< ValueT > & d_values
 [in,out] Double-buffer whose current buffer contains the unsorted input values and, upon return, is updated to point to the sorted output values
 
OffsetT num_items
 [in] Number of items to sort
 
int begin_bit
 [in] The beginning (least-significant) bit index needed for key comparison
 
int end_bit
 [in] The past-the-end (most-significant) bit index needed for key comparison
 
cudaStream_t stream
 [in] CUDA stream to launch kernels within. Default is stream0.
 
bool debug_synchronous
 [in] Whether 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.
 
int ptx_version
 [in] PTX version
 
bool is_overwrite_okay
 [in] Whether is okay to overwrite source buffers
 

Member Enumeration Documentation

◆ anonymous enum

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
anonymous enum

Definition at line 814 of file dispatch_radix_sort.cuh.

Constructor & Destructor Documentation

◆ DispatchRadixSort()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
CUB_RUNTIME_FUNCTION __forceinline__ cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::DispatchRadixSort ( void *  d_temp_storage,
size_t &  temp_storage_bytes,
DoubleBuffer< KeyT > &  d_keys,
DoubleBuffer< ValueT > &  d_values,
OffsetT  num_items,
int  begin_bit,
int  end_bit,
bool  is_overwrite_okay,
cudaStream_t  stream,
bool  debug_synchronous,
int  ptx_version 
)
inline

Constructor.

Definition at line 844 of file dispatch_radix_sort.cuh.

Member Function Documentation

◆ Dispatch()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::Dispatch ( void *  d_temp_storage,
size_t &  temp_storage_bytes,
DoubleBuffer< KeyT > &  d_keys,
DoubleBuffer< ValueT > &  d_values,
OffsetT  num_items,
int  begin_bit,
int  end_bit,
bool  is_overwrite_okay,
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,out]d_keysDouble-buffer whose current buffer contains the unsorted input keys and, upon return, is updated to point to the sorted output keys
[in,out]d_valuesDouble-buffer whose current buffer contains the unsorted input values and, upon return, is updated to point to the sorted output values
[in]num_itemsNumber of items to sort
[in]begin_bitThe beginning (least-significant) bit index needed for key comparison
[in]end_bitThe past-the-end (most-significant) bit index needed for key comparison
[in]is_overwrite_okayWhether is okay to overwrite source buffers
[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 1255 of file dispatch_radix_sort.cuh.

◆ Invoke()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
template<typename ActivePolicyT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::Invoke ( )
inline

Invocation.

Definition at line 1222 of file dispatch_radix_sort.cuh.

◆ InvokePass()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
template<typename PassConfigT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::InvokePass ( const KeyT *  d_keys_in,
KeyT *  d_keys_out,
const ValueT *  d_values_in,
ValueT *  d_values_out,
OffsetT *  d_spine,
int  spine_length,
int &  current_bit,
PassConfigT &  pass_config 
)
inline

Invoke a three-kernel sorting pass at the current bit.

Definition at line 945 of file dispatch_radix_sort.cuh.

◆ InvokePasses()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
template<typename ActivePolicyT , typename UpsweepKernelT , typename ScanKernelT , typename DownsweepKernelT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::InvokePasses ( UpsweepKernelT  upsweep_kernel,
UpsweepKernelT  alt_upsweep_kernel,
ScanKernelT  scan_kernel,
DownsweepKernelT  downsweep_kernel,
DownsweepKernelT  alt_downsweep_kernel 
)
inline

Invocation (run multiple digit passes)

< Function type of cub::DeviceRadixSortDownsweepKernel

Parameters
[in]upsweep_kernelKernel function pointer to parameterization of cub::DeviceRadixSortUpsweepKernel
[in]alt_upsweep_kernelAlternate kernel function pointer to parameterization of cub::DeviceRadixSortUpsweepKernel
[in]scan_kernelKernel function pointer to parameterization of cub::SpineScanKernel
[in]downsweep_kernelKernel function pointer to parameterization of cub::DeviceRadixSortDownsweepKernel
[in]alt_downsweep_kernelAlternate kernel function pointer to parameterization of cub::DeviceRadixSortDownsweepKernel

Definition at line 1096 of file dispatch_radix_sort.cuh.

◆ InvokeSingleTile()

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
template<typename ActivePolicyT , typename SingleTileKernelT >
CUB_RUNTIME_FUNCTION __forceinline__ cudaError_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::InvokeSingleTile ( SingleTileKernelT  single_tile_kernel)
inline

Invoke a single block to sort in-core.

< Function type of cub::DeviceRadixSortSingleTileKernel

Parameters
[in]single_tile_kernelKernel function pointer to parameterization of cub::DeviceRadixSortSingleTileKernel

Definition at line 880 of file dispatch_radix_sort.cuh.

Field Documentation

◆ begin_bit

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
int cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::begin_bit

[in] The beginning (least-significant) bit index needed for key comparison

Definition at line 830 of file dispatch_radix_sort.cuh.

◆ d_keys

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
DoubleBuffer<KeyT>& cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::d_keys

[in,out] Double-buffer whose current buffer contains the unsorted input keys and, upon return, is updated to point to the sorted output keys

Definition at line 827 of file dispatch_radix_sort.cuh.

◆ d_temp_storage

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
void* cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::d_temp_storage

[in] Device-accessible allocation of temporary storage. When NULL, the required allocation size is written to temp_storage_bytes and no work is done.

Definition at line 825 of file dispatch_radix_sort.cuh.

◆ d_values

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
DoubleBuffer<ValueT>& cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::d_values

[in,out] Double-buffer whose current buffer contains the unsorted input values and, upon return, is updated to point to the sorted output values

Definition at line 828 of file dispatch_radix_sort.cuh.

◆ debug_synchronous

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
bool cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::debug_synchronous

[in] Whether 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 833 of file dispatch_radix_sort.cuh.

◆ end_bit

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
int cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::end_bit

[in] The past-the-end (most-significant) bit index needed for key comparison

Definition at line 831 of file dispatch_radix_sort.cuh.

◆ is_overwrite_okay

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
bool cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::is_overwrite_okay

[in] Whether is okay to overwrite source buffers

Definition at line 835 of file dispatch_radix_sort.cuh.

◆ num_items

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
OffsetT cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::num_items

[in] Number of items to sort

Definition at line 829 of file dispatch_radix_sort.cuh.

◆ ptx_version

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
int cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::ptx_version

[in] PTX version

Definition at line 834 of file dispatch_radix_sort.cuh.

◆ stream

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
cudaStream_t cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::stream

[in] CUDA stream to launch kernels within. Default is stream0.

Definition at line 832 of file dispatch_radix_sort.cuh.

◆ temp_storage_bytes

template<bool IS_DESCENDING, typename KeyT , typename ValueT , typename OffsetT >
size_t& cub::DispatchRadixSort< IS_DESCENDING, KeyT, ValueT, OffsetT >::temp_storage_bytes

[in,out] Reference to size in bytes of d_temp_storage allocation

Definition at line 826 of file dispatch_radix_sort.cuh.

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