OpenFPM_pdata  4.1.0
Project that contain the implementation of distributed structures
cub::ReduceByKeyScanTileState< ValueT, KeyT, true > Struct Template Reference

Detailed Description

template<typename ValueT, typename KeyT>
struct cub::ReduceByKeyScanTileState< ValueT, KeyT, true >

Tile status interface for reduction by key, specialized for scan status and value types that can be combined into one machine word that can be read/written coherently in a single access.

Definition at line 478 of file single_pass_scan_operators.cuh.

Data Structures

struct  TileDescriptorBigStatus
 
struct  TileDescriptorLittleStatus
 

Public Types

enum  { PAIR_SIZE = sizeof(ValueT) + sizeof(KeyT), TXN_WORD_SIZE = 1 << Log2<PAIR_SIZE + 1>::VALUE, STATUS_WORD_SIZE = TXN_WORD_SIZE - PAIR_SIZE, TILE_STATUS_PADDING = CUB_PTX_WARP_THREADS }
 
typedef KeyValuePair< KeyT, ValueT > KeyValuePairT
 
typedef If<(STATUS_WORD_SIZE==8), long long, typename If<(STATUS_WORD_SIZE==4), int, typename If<(STATUS_WORD_SIZE==2), short, char >::Type >::Type >::Type StatusWord
 
typedef If<(TXN_WORD_SIZE==16), longlong2, typename If<(TXN_WORD_SIZE==8), long long, int >::Type >::Type TxnWord
 
typedef If<(sizeof(ValueT)==sizeof(KeyT)), TileDescriptorBigStatus, TileDescriptorLittleStatus >::Type TileDescriptor
 

Public Member Functions

__host__ __device__ __forceinline__ ReduceByKeyScanTileState ()
 Constructor.
 
__host__ __device__ __forceinline__ cudaError_t Init (int, void *d_temp_storage, size_t)
 Initializer. More...
 
__device__ __forceinline__ void InitializeStatus (int num_tiles)
 
__device__ __forceinline__ void SetInclusive (int tile_idx, KeyValuePairT tile_inclusive)
 
__device__ __forceinline__ void SetPartial (int tile_idx, KeyValuePairT tile_partial)
 
__device__ __forceinline__ void WaitForValid (int tile_idx, StatusWord &status, KeyValuePairT &value)
 

Static Public Member Functions

__host__ __device__ static __forceinline__ cudaError_t AllocationSize (int num_tiles, size_t &temp_storage_bytes)
 

Data Fields

TxnWord * d_tile_descriptors
 

Member Function Documentation

◆ AllocationSize()

template<typename ValueT , typename KeyT >
__host__ __device__ static __forceinline__ cudaError_t cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::AllocationSize ( int  num_tiles,
size_t &  temp_storage_bytes 
)
inlinestatic

Compute device memory needed for tile status

Parameters
[in]num_tilesNumber of tiles
[out]temp_storage_bytesSize in bytes of \t d_temp_storage allocation

Definition at line 560 of file single_pass_scan_operators.cuh.

◆ Init()

template<typename ValueT , typename KeyT >
__host__ __device__ __forceinline__ cudaError_t cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::Init ( int  ,
void *  d_temp_storage,
size_t   
)
inline

Initializer.

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.

Definition at line 546 of file single_pass_scan_operators.cuh.

◆ InitializeStatus()

template<typename ValueT , typename KeyT >
__device__ __forceinline__ void cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::InitializeStatus ( int  num_tiles)
inline

Initialize (from device)

Definition at line 572 of file single_pass_scan_operators.cuh.

◆ SetInclusive()

template<typename ValueT , typename KeyT >
__device__ __forceinline__ void cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::SetInclusive ( int  tile_idx,
KeyValuePairT  tile_inclusive 
)
inline

Update the specified tile's inclusive value and corresponding status

Definition at line 597 of file single_pass_scan_operators.cuh.

◆ SetPartial()

template<typename ValueT , typename KeyT >
__device__ __forceinline__ void cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::SetPartial ( int  tile_idx,
KeyValuePairT  tile_partial 
)
inline

Update the specified tile's partial value and corresponding status

Definition at line 613 of file single_pass_scan_operators.cuh.

◆ WaitForValid()

template<typename ValueT , typename KeyT >
__device__ __forceinline__ void cub::ReduceByKeyScanTileState< ValueT, KeyT, true >::WaitForValid ( int  tile_idx,
StatusWord &  status,
KeyValuePairT value 
)
inline

Wait for the corresponding tile to become non-invalid

Definition at line 628 of file single_pass_scan_operators.cuh.


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