Detailed Description

Functions
template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >
__global__ void	nvbio::cuda::wq::inplace_work_queue_kernel (const WorkStreamT stream, WorkQueueStats::View stats)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT , bool DO_LOADS>
__global__ void	nvbio::cuda::wq::mk_work_queue_kernel (const uint32 n_tile_grids, typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context context, const uint32 in_queue_id, const uint32 in_queue_size, const WorkStreamT stream, uint32 stream_begin)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >
__global__ void	nvbio::cuda::wq::mk_load_kernel (const uint32 n_tile_grids, typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context context, const uint32 in_queue_id, const uint32 in_queue_size, const WorkStreamT stream, uint32 stream_begin)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStream , typename WorkMover >
__global__ void	nvbio::cuda::wq::mk_move_kernel (const uint32 n_tile_grids, typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context context, const uint32 in_queue_id, const uint32 in_queue_size, const WorkStream stream, const WorkMover mover)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT , typename WorkMoverT >
__global__ void	nvbio::cuda::wq::work_queue_kernel (const uint32 n_tile_grids, typename WorkQueue< OrderedQueueTag, WorkUnitT, BLOCKDIM >::Context context, const WorkStreamT stream, const WorkMoverT mover)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >
__global__ void	nvbio::cuda::wq::persistent_warps_work_queue_kernel (uint32 *pool, const WorkStreamT stream, WorkQueueStats::View stats)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >
__global__ void	nvbio::cuda::wq::__launch_bounds__ (BLOCKDIM, 6) persistent_threads_work_queue_kernel(uint32 *pool

	nvbio::cuda::wq::while (__any(active))

Variables
__global__ void const uint32	nvbio::cuda::wq::max_inactive_lanes

__global__ void const uint32 const WorkStreamT	nvbio::cuda::wq::stream

__global__ void const uint32 const WorkStreamT WorkQueueStats::View	nvbio::cuda::wq::stats

const uint32	nvbio::cuda::wq::thread_id = threadIdx.x + blockIdx.x*BLOCKDIM

WorkUnit	nvbio::cuda::wq::unit

const uint32	nvbio::cuda::wq::NUM_WARPS = BLOCKDIM >> cuda::Arch::LOG_WARP_SIZE

__shared__ volatile uint32	nvbio::cuda::wq::sm_broadcast [NUM_WARPS]

const uint32	nvbio::cuda::wq::invalid_unit = uint32(-1)

uint32	nvbio::cuda::wq::work_id = invalid_unit

uint32	nvbio::cuda::wq::work_iter = 0u

bool	nvbio::cuda::wq::active = true

const uint32	nvbio::cuda::wq::stream_size = stream.size()

Function Documentation

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >

__global__ void nvbio::cuda::wq::__launch_bounds__	(	BLOCKDIM	,
		6
	)

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >

__global__ void nvbio::cuda::wq::inplace_work_queue_kernel	(	const WorkStreamT	stream,
		WorkQueueStats::View	stats
	)

Definition at line 52 of file work_queue_inl.h.

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >

__global__ void nvbio::cuda::wq::mk_load_kernel	(	const uint32	n_tile_grids,
		typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context	context,
		const uint32	in_queue_id,
		const uint32	in_queue_size,
		const WorkStreamT	stream,
		uint32	stream_begin
	)

Definition at line 122 of file work_queue_multipass_inl.h.

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStream , typename WorkMover >

__global__ void nvbio::cuda::wq::mk_move_kernel	(	const uint32	n_tile_grids,
		typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context	context,
		const uint32	in_queue_id,
		const uint32	in_queue_size,
		const WorkStream	stream,
		const WorkMover	mover
	)

Definition at line 172 of file work_queue_multipass_inl.h.

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT , bool DO_LOADS>

__global__ void nvbio::cuda::wq::mk_work_queue_kernel	(	const uint32	n_tile_grids,
		typename WorkQueue< MultiPassQueueTag, WorkUnitT, BLOCKDIM >::Context	context,
		const uint32	in_queue_id,
		const uint32	in_queue_size,
		const WorkStreamT	stream,
		uint32	stream_begin
	)

Definition at line 53 of file work_queue_multipass_inl.h.

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT >

__global__ void nvbio::cuda::wq::persistent_warps_work_queue_kernel	(	uint32 *	pool,
		const WorkStreamT	stream,
		WorkQueueStats::View	stats
	)

Definition at line 56 of file work_queue_persistent_inl.h.

nvbio::cuda::wq::while ( __any(active) )

Definition at line 165 of file work_queue_persistent_inl.h.

template<uint32 BLOCKDIM, typename WorkUnitT , typename WorkStreamT , typename WorkMoverT >

__global__ void nvbio::cuda::wq::work_queue_kernel	(	const uint32	n_tile_grids,
		typename WorkQueue< OrderedQueueTag, WorkUnitT, BLOCKDIM >::Context	context,
		const WorkStreamT	stream,
		const WorkMoverT	mover
	)