cub::DispatchSpmv< ValueT, OffsetT > Struct Template Reference

< Signed integer type for global offsets More...

Detailed Description

template<typename ValueT, typename OffsetT>
struct cub::DispatchSpmv< ValueT, OffsetT >

< Signed integer type for global offsets

Utility class for dispatching the appropriately-tuned kernels for DeviceSpmv

Definition at line 237 of file dispatch_spmv_orig.cuh.

Data Structures
struct	KernelConfig
struct	Policy110
	SM11. More...
struct	Policy200
	SM20. More...
struct	Policy300
	SM30. More...
struct	Policy350
	SM35. More...
struct	Policy370
	SM37. More...
struct	Policy500
	SM50. More...
struct	Policy600
	SM60. More...
struct	PtxSegmentFixupPolicy
struct	PtxSpmvPolicyT

Public Types
enum	{ INIT_KERNEL_THREADS = 128 }
typedef SpmvParams< ValueT, OffsetT >	SpmvParamsT
typedef CubVector< OffsetT, 2 >::Type	CoordinateT
typedef ReduceByKeyScanTileState< ValueT, OffsetT >	ScanTileStateT
typedef KeyValuePair< OffsetT, ValueT >	KeyValuePairT
typedef Policy110	PtxPolicy

Static Public Member Functions
template<typename KernelConfig >
CUB_RUNTIME_FUNCTION static __forceinline__ void	InitConfigs (int ptx_version, KernelConfig &spmv_config, KernelConfig &segment_fixup_config)
template<typename Spmv1ColKernelT , typename SpmvSearchKernelT , typename SpmvKernelT , typename SegmentFixupKernelT >
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t	Dispatch (void *d_temp_storage, size_t &temp_storage_bytes, SpmvParamsT &spmv_params, cudaStream_t stream, bool debug_synchronous, Spmv1ColKernelT spmv_1col_kernel, SpmvSearchKernelT spmv_search_kernel, SpmvKernelT spmv_kernel, SegmentFixupKernelT segment_fixup_kernel, KernelConfig spmv_config, KernelConfig segment_fixup_config)
	< Function type of cub::DeviceSegmentFixupKernelT
CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t	Dispatch (void *d_temp_storage, size_t &temp_storage_bytes, SpmvParamsT &spmv_params, cudaStream_t stream=0, bool debug_synchronous=false)

Member Typedef Documentation

CoordinateT

template<typename ValueT , typename OffsetT >

typedef CubVector<OffsetT,2>::Type cub::DispatchSpmv< ValueT, OffsetT >::CoordinateT

Definition at line 252 of file dispatch_spmv_orig.cuh.

KeyValuePairT

template<typename ValueT , typename OffsetT >

typedef KeyValuePair<OffsetT, ValueT> cub::DispatchSpmv< ValueT, OffsetT >::KeyValuePairT

Definition at line 258 of file dispatch_spmv_orig.cuh.

PtxPolicy

template<typename ValueT , typename OffsetT >

typedef Policy110 cub::DispatchSpmv< ValueT, OffsetT >::PtxPolicy

Definition at line 468 of file dispatch_spmv_orig.cuh.

ScanTileStateT

template<typename ValueT , typename OffsetT >

typedef ReduceByKeyScanTileState<ValueT, OffsetT> cub::DispatchSpmv< ValueT, OffsetT >::ScanTileStateT

Definition at line 255 of file dispatch_spmv_orig.cuh.

SpmvParamsT

template<typename ValueT , typename OffsetT >

typedef SpmvParams<ValueT, OffsetT> cub::DispatchSpmv< ValueT, OffsetT >::SpmvParamsT

Definition at line 249 of file dispatch_spmv_orig.cuh.

Member Enumeration Documentation

anonymous enum

template<typename ValueT , typename OffsetT >

anonymous enum

Definition at line 243 of file dispatch_spmv_orig.cuh.

Member Function Documentation

Dispatch() [1/2]

template<typename ValueT , typename OffsetT >

template<typename Spmv1ColKernelT , typename SpmvSearchKernelT , typename SpmvKernelT , typename SegmentFixupKernelT >

CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t cub::DispatchSpmv< ValueT, OffsetT >::Dispatch ( void *  d_temp_storage, size_t &  temp_storage_bytes, SpmvParamsT &  spmv_params, cudaStream_t  stream, bool  debug_synchronous, Spmv1ColKernelT  spmv_1col_kernel, SpmvSearchKernelT  spmv_search_kernel, SpmvKernelT  spmv_kernel, SegmentFixupKernelT  segment_fixup_kernel, KernelConfig  spmv_config, KernelConfig  segment_fixup_config  )

inlinestatic

< Function type of cub::DeviceSegmentFixupKernelT

Internal dispatch routine for computing a device-wide reduction using the specified kernel functions.

If the input is larger than a single tile, this method uses two-passes of kernel invocations.

Parameters

[in]	d_temp_storage	Device-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_bytes	Reference to size in bytes of d_temp_storage allocation
	spmv_params	SpMV input parameter bundle
[in]	stream	CUDA stream to launch kernels within. Default is stream0.
[in]	debug_synchronous	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.
[in]	spmv_1col_kernel	Kernel function pointer to parameterization of DeviceSpmv1ColKernel
[in]	spmv_search_kernel	Kernel function pointer to parameterization of AgentSpmvSearchKernel
[in]	spmv_kernel	Kernel function pointer to parameterization of AgentSpmvKernel
[in]	segment_fixup_kernel	Kernel function pointer to parameterization of cub::DeviceSegmentFixupKernel
[in]	spmv_config	Dispatch parameters that match the policy that spmv_kernel was compiled for
[in]	segment_fixup_config	Dispatch parameters that match the policy that segment_fixup_kernel was compiled for

Definition at line 578 of file dispatch_spmv_orig.cuh.

Dispatch() [2/2]

template<typename ValueT , typename OffsetT >

CUB_RUNTIME_FUNCTION static __forceinline__ cudaError_t cub::DispatchSpmv< ValueT, OffsetT >::Dispatch ( void *  d_temp_storage, size_t &  temp_storage_bytes, SpmvParamsT &  spmv_params, cudaStream_t  stream = 0, bool  debug_synchronous = false  )

inlinestatic

Internal dispatch routine for computing a device-wide reduction

Parameters

[in]	d_temp_storage	Device-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_bytes	Reference to size in bytes of d_temp_storage allocation
	spmv_params	SpMV input parameter bundle
[in]	stream	[optional] CUDA stream to launch kernels within. Default is stream0.
[in]	debug_synchronous	[optional] Whether or not to synchronize the stream after every kernel launch to check for errors. May cause significant slowdown. Default is false.

Definition at line 793 of file dispatch_spmv_orig.cuh.

InitConfigs()

template<typename ValueT , typename OffsetT >

template<typename KernelConfig >

CUB_RUNTIME_FUNCTION static __forceinline__ void cub::DispatchSpmv< ValueT, OffsetT >::InitConfigs ( int  ptx_version, KernelConfig &  spmv_config, KernelConfig &  segment_fixup_config  )

inlinestatic

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

Definition at line 486 of file dispatch_spmv_orig.cuh.

---

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

• openfpm_data/src/util/cuda/cub_old/device/dispatch/dispatch_spmv_orig.cuh