Thread-Level Operations (Utilities)#
Scan#
exclusive#
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_exclusive(T (&input)[LENGTH], T (&output)[LENGTH], ScanOp scan_op, T prefix, bool apply_prefix = true)# Perform a sequential exclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
prefix – [in] Prefix to seed scan with
apply_prefix – [in] Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
prefix – [in] - Prefix to seed scan with
apply_prefix – [in] - Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
- Template Parameters:
LENGTH – - Length of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_exclusive(T *input, T *output, ScanOp scan_op, T prefix, bool apply_prefix = true)# Perform a sequential exclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
prefix – [in] Prefix to seed scan with
apply_prefix – [in] Whether or not the calling thread should apply its prefix. If not, the first output element is undefined. (Handy for preventing thread-0 from applying a prefix.)
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
prefix – [in] - Prefix to seed scan with
apply_prefix – [in] - Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
- Template Parameters:
LENGTH – - Length of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_exclusive(T inclusive, T exclusive, T *input, T *output, ScanOp scan_op, Int2Type<LENGTH>)# Perform a sequential exclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
inclusive – [in] Initial value for inclusive aggregate
exclusive – [in] Initial value for exclusive aggregate
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
inclusive – [in] - Initial value for inclusive aggregate
exclusive – [in] - Initial value for exclusive aggregate
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
- Template Parameters:
LENGTH – - Length of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
inclusive#
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_inclusive(T inclusive, T *input, T *output, ScanOp scan_op, Int2Type<LENGTH>)# Perform a sequential exclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
inclusive – [in] Initial value for inclusive aggregate
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
inclusive – [in] - Initial value for inclusive aggregate
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
- Template Parameters:
LENGTH – - Length of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_inclusive(T *input, T *output, ScanOp scan_op)# Perform a sequential inclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Template Parameters:
LENGTH – - LengthT of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Parameters:
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_inclusive(T (&input)[LENGTH], T (&output)[LENGTH], ScanOp scan_op)# Perform a sequential inclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
- Template Parameters:
LENGTH – - LengthT of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_inclusive(T *input, T *output, ScanOp scan_op, T prefix, bool apply_prefix = true)# Perform a sequential inclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Parameters:
input – [in] Input array
output – [out] Output array (may be aliased to
input
)scan_op – [in] Binary scan operator
prefix – [in] Prefix to seed scan with
apply_prefix – [in] Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
prefix – [in] - Prefix to seed scan with
apply_prefix – [in] - Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
- Template Parameters:
LENGTH – - LengthT of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Returns:
- Aggregate of the scan
-
template<int LENGTH, typename T, typename ScanOp>
__device__ inline T rocprim::thread_scan_inclusive(T (&input)[LENGTH], T (&output)[LENGTH], ScanOp scan_op, T prefix, bool apply_prefix = true)# Perform a sequential inclusive prefix scan over
LENGTH
elements of theinput
array. The aggregate is returned.- Template Parameters:
LENGTH – - LengthT of
input
andoutput
arraysT – - [inferred] The data type to be scanned.
ScanOp – - [inferred] Binary scan operator type having member
T operator()(const T &a, const T &b)
- Parameters:
input – [in] - Input array
output – [out] - Output array (may be aliased to
input
)scan_op – [in] - Binary scan operator
prefix – [in] - Prefix to seed scan with
apply_prefix – [in] - Whether or not the calling thread should apply its prefix. (Handy for preventing thread-0 from applying a prefix.)
- Returns:
- Aggregate of the scan