SegmentedScanTraits< T, oper, backward, exclusivity, doShiftFlags, fullBlock, sums, sm12OrBetter > Class Template Reference

Template class containing compile-time parameters to the segmented scan functions. More...

List of all members.

Static Public Member Functions
static __device__ bool	isBackward ()
	Returns true if this is a backward scan.
static __device__ bool	isExclusive ()
	Returns true if this is an exclusive scan.
static __device__ bool	shiftFlags ()
static __device__ bool	isFullBlock ()
	Returns true if this is a full scan -- all blocks process CTA_SIZE * SCAN_ELEMENTS_PER_THREAD elements.
static __device__ bool	writeSums ()
	Returns true if this scan writes the sum of each block to the d_blockSums array (multi-block scans)
static __device__ bool	isSM12OrBetter ()
	Returns true if we are sm12 or better hardware.
static __device__ T	op (const T &a, const T &b)
	The operator function used for segmented scan.
static __device__ T	identity ()
	The identity value used by segmented scan.

---

Detailed Description

template<class T, CUDPPOperator oper, bool backward, bool exclusivity, bool doShiftFlags, bool fullBlock, bool sums, bool sm12OrBetter>
class SegmentedScanTraits< T, oper, backward, exclusivity, doShiftFlags, fullBlock, sums, sm12OrBetter >

Template class containing compile-time parameters to the segmented scan functions.

SegmentedScanTraits is passed as a template parameter to all segmented scan functions. By using these compile-time functions we can enable generic code while maintaining the highest performance. This is crucial for the performance of low-level workhorse algorithms like segmented scan.

Parameters:

T	The datatype of the segmented scan
oper	The CUDPPOperator to use for the segmented scan (add, max, etc.)
unroll	True if scan inner loops should be unrolled
sums	True if each block should write it's sum to the d_blockSums array (false for single-block scans)
backward	True if this is a backward scan, False if this is a forward scan
fullBlock	True if all blocks in this scan are full (CTA_SIZE * SCAN_ELEMENTS_PER_THREAD elements)
exclusivity	True for exclusive scans, false for inclusive scans

---

Member Function Documentation

template<class T , CUDPPOperator oper, bool backward, bool exclusivity, bool doShiftFlags, bool fullBlock, bool sums, bool sm12OrBetter>

static __device__ bool SegmentedScanTraits< T, oper, backward, exclusivity, doShiftFlags, fullBlock, sums, sm12OrBetter >::shiftFlags

(

)

[inline, static]

Returns true if this scan needs to shift flags to the left. This is only needed for the first level scan in a multi-block scan

---

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

• segmented_scan_cta.cu