cv::gpu::device::reduce_key_val_detail Namespace Reference

Classes
struct	Dispatcher
struct	For
struct	For< N, N >
struct	Generic
struct	GenericOptimized32
struct	GetType
struct	GetType< T & >
struct	GetType< T * >
struct	GetType< volatile T * >
struct	IsPowerOf2
struct	StaticIf
struct	StaticIf< false, T1, T2 >
struct	StaticIf< true, T1, T2 >
struct	Unroll
struct	Unroll< 0, KP, KR, VP, VR, Cmp >
struct	WarpOptimized

Functions
template<typename T >
__device__ __forceinline__ void	loadToSmem (volatile T *smem, T &data, unsigned int tid)
template<typename T >
__device__ __forceinline__ void	loadFromSmem (volatile T *smem, T &data, unsigned int tid)
template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >
__device__ __forceinline__ void	loadToSmem (const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &smem, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &data, unsigned int tid)
template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >
__device__ __forceinline__ void	loadFromSmem (const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &smem, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &data, unsigned int tid)
template<typename V >
__device__ __forceinline__ void	copyValsShfl (V &val, unsigned int delta, int width)
template<typename V >
__device__ __forceinline__ void	copyVals (volatile V *svals, V &val, unsigned int tid, unsigned int delta)
template<typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >
__device__ __forceinline__ void	copyValsShfl (const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, unsigned int delta, int width)
template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >
__device__ __forceinline__ void	copyVals (const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &svals, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, unsigned int tid, unsigned int delta)
template<typename K , typename V , class Cmp >
__device__ __forceinline__ void	mergeShfl (K &key, V &val, const Cmp &cmp, unsigned int delta, int width)
template<typename K , typename V , class Cmp >
__device__ __forceinline__ void	merge (volatile K *skeys, K &key, volatile V *svals, V &val, const Cmp &cmp, unsigned int tid, unsigned int delta)
template<typename K , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp >
__device__ __forceinline__ void	mergeShfl (K &key, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, const Cmp &cmp, unsigned int delta, int width)
template<typename K , typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp >
__device__ __forceinline__ void	merge (volatile K *skeys, K &key, const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &svals, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, const Cmp &cmp, unsigned int tid, unsigned int delta)
template<typename KR0 , typename KR1 , typename KR2 , typename KR3 , typename KR4 , typename KR5 , typename KR6 , typename KR7 , typename KR8 , typename KR9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp0 , class Cmp1 , class Cmp2 , class Cmp3 , class Cmp4 , class Cmp5 , class Cmp6 , class Cmp7 , class Cmp8 , class Cmp9 >
__device__ __forceinline__ void	mergeShfl (const thrust::tuple< KR0, KR1, KR2, KR3, KR4, KR5, KR6, KR7, KR8, KR9 > &key, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, const thrust::tuple< Cmp0, Cmp1, Cmp2, Cmp3, Cmp4, Cmp5, Cmp6, Cmp7, Cmp8, Cmp9 > &cmp, unsigned int delta, int width)
template<typename KP0 , typename KP1 , typename KP2 , typename KP3 , typename KP4 , typename KP5 , typename KP6 , typename KP7 , typename KP8 , typename KP9 , typename KR0 , typename KR1 , typename KR2 , typename KR3 , typename KR4 , typename KR5 , typename KR6 , typename KR7 , typename KR8 , typename KR9 , typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp0 , class Cmp1 , class Cmp2 , class Cmp3 , class Cmp4 , class Cmp5 , class Cmp6 , class Cmp7 , class Cmp8 , class Cmp9 >
__device__ __forceinline__ void	merge (const thrust::tuple< KP0, KP1, KP2, KP3, KP4, KP5, KP6, KP7, KP8, KP9 > &skeys, const thrust::tuple< KR0, KR1, KR2, KR3, KR4, KR5, KR6, KR7, KR8, KR9 > &key, const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &svals, const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &val, const thrust::tuple< Cmp0, Cmp1, Cmp2, Cmp3, Cmp4, Cmp5, Cmp6, Cmp7, Cmp8, Cmp9 > &cmp, unsigned int tid, unsigned int delta)

Function Documentation

template<typename V >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::copyVals	(	volatile V *	svals,
		V &	val,
		unsigned int	tid,
		unsigned int	delta
	)

template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::copyVals	(	const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &	svals,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		unsigned int	tid,
		unsigned int	delta
	)

template<typename V >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::copyValsShfl	(	V &	val,
		unsigned int	delta,
		int	width
	)

template<typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::copyValsShfl	(	const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		unsigned int	delta,
		int	width
	)

template<typename T >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::loadFromSmem	(	volatile T *	smem,
		T &	data,
		unsigned int	tid
	)

template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::loadFromSmem	(	const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &	smem,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	data,
		unsigned int	tid
	)

template<typename T >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::loadToSmem	(	volatile T *	smem,
		T &	data,
		unsigned int	tid
	)

template<typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::loadToSmem	(	const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &	smem,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	data,
		unsigned int	tid
	)

template<typename K , typename V , class Cmp >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::merge	(	volatile K *	skeys,
		K &	key,
		volatile V *	svals,
		V &	val,
		const Cmp &	cmp,
		unsigned int	tid,
		unsigned int	delta
	)

template<typename K , typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::merge	(	volatile K *	skeys,
		K &	key,
		const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &	svals,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		const Cmp &	cmp,
		unsigned int	tid,
		unsigned int	delta
	)

template<typename KP0 , typename KP1 , typename KP2 , typename KP3 , typename KP4 , typename KP5 , typename KP6 , typename KP7 , typename KP8 , typename KP9 , typename KR0 , typename KR1 , typename KR2 , typename KR3 , typename KR4 , typename KR5 , typename KR6 , typename KR7 , typename KR8 , typename KR9 , typename VP0 , typename VP1 , typename VP2 , typename VP3 , typename VP4 , typename VP5 , typename VP6 , typename VP7 , typename VP8 , typename VP9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp0 , class Cmp1 , class Cmp2 , class Cmp3 , class Cmp4 , class Cmp5 , class Cmp6 , class Cmp7 , class Cmp8 , class Cmp9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::merge	(	const thrust::tuple< KP0, KP1, KP2, KP3, KP4, KP5, KP6, KP7, KP8, KP9 > &	skeys,
		const thrust::tuple< KR0, KR1, KR2, KR3, KR4, KR5, KR6, KR7, KR8, KR9 > &	key,
		const thrust::tuple< VP0, VP1, VP2, VP3, VP4, VP5, VP6, VP7, VP8, VP9 > &	svals,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		const thrust::tuple< Cmp0, Cmp1, Cmp2, Cmp3, Cmp4, Cmp5, Cmp6, Cmp7, Cmp8, Cmp9 > &	cmp,
		unsigned int	tid,
		unsigned int	delta
	)

template<typename K , typename V , class Cmp >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::mergeShfl	(	K &	key,
		V &	val,
		const Cmp &	cmp,
		unsigned int	delta,
		int	width
	)

template<typename K , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::mergeShfl	(	K &	key,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		const Cmp &	cmp,
		unsigned int	delta,
		int	width
	)

template<typename KR0 , typename KR1 , typename KR2 , typename KR3 , typename KR4 , typename KR5 , typename KR6 , typename KR7 , typename KR8 , typename KR9 , typename VR0 , typename VR1 , typename VR2 , typename VR3 , typename VR4 , typename VR5 , typename VR6 , typename VR7 , typename VR8 , typename VR9 , class Cmp0 , class Cmp1 , class Cmp2 , class Cmp3 , class Cmp4 , class Cmp5 , class Cmp6 , class Cmp7 , class Cmp8 , class Cmp9 >

__device__ __forceinline__ void cv::gpu::device::reduce_key_val_detail::mergeShfl	(	const thrust::tuple< KR0, KR1, KR2, KR3, KR4, KR5, KR6, KR7, KR8, KR9 > &	key,
		const thrust::tuple< VR0, VR1, VR2, VR3, VR4, VR5, VR6, VR7, VR8, VR9 > &	val,
		const thrust::tuple< Cmp0, Cmp1, Cmp2, Cmp3, Cmp4, Cmp5, Cmp6, Cmp7, Cmp8, Cmp9 > &	cmp,
		unsigned int	delta,
		int	width
	)