Sort.cl.in File Reference

Bitonic sort OpenCL methods. (See Aqua::CalcServer::Sort for details) More...

Macros
#define	SWAP(address)
#define	SORT(address)

Functions
__kernel void	init (const __global T *restrict in_vals, __global T *restrict vals, __global usize *keys, usize n, usize N)
	Initializes the padded values and keys (permutations)
	SWAP (__local)
	SORT (__local)
	SWAP (__global)
	SORT (__global)
__kernel void	bitonic_start (__global T *restrict vals, __global usize *restrict keys)
	Do the first half of the sort & merge stages.
__kernel void	bitonic_local (__global T *restrict vals, __global usize *restrict keys, const usize N, const usize bsize, usize stride)
	Do a single sort & merge step.
__kernel void	bitonic_global (__global T *restrict vals, __global usize *restrict keys, const usize N, const usize bsize, const usize stride)
	Do a single sort & merge step.
__kernel void	inverse_keys (const __global usize *restrict keys, __global usize *inv_keys, usize n)

Detailed Description

Bitonic sort OpenCL methods. (See Aqua::CalcServer::Sort for details)

Note

The header CalcServer/Sort.hcl.in is automatically appended.

Macro Definition Documentation

SORT

#define SORT

(

address

)

Value:

inline void                                                                    \
sort ## address(address T *a,                                                  \
                address T *b,                                                  \
                address uint *ka,                                              \
                address uint *kb,                                              \
                const char dir)                                                \
{                                                                              \
        if ((*a > *b) == dir) swap ## address(a, b, ka, kb);                       \
}

SWAP

#define SWAP

(

address

)

Value:

inline void                                                                    \
swap ## address(address T *a,                                                  \
                address T *b,                                                  \
                address uint *ka,                                              \
                address uint *kb)                                              \
{                                                                              \
        T tmp;                                                                     \
        tmp = *b; *b = *a; *a = tmp;                                               \
        uint ktmp;                                                                 \
        ktmp = *kb; *kb = *ka; *ka = ktmp;                                         \
}

Function Documentation

bitonic_global()

__kernel void bitonic_global	(	__global T *restrict	vals,
		__global usize *restrict	keys,
		const usize	N,
		const usize	bsize,
		const usize	stride )

Do a single sort & merge step.

Use this function when the operation does not fit on the local memory

Parameters

vals	Values
keys	Keys (permutations)
N	Padded number of values.
bsize	Sorting block size
stride	Step length within the block

bitonic_local()

__kernel void bitonic_local	(	__global T *restrict	vals,
		__global usize *restrict	keys,
		const usize	N,
		const usize	bsize,
		usize	stride )

Do a single sort & merge step.

Use this function when the operation still fits on the local memory

Parameters

vals	Values
keys	Keys (permutations)
N	Padded number of values.
bsize	Sorting block size
stride	Step length within the block

bitonic_start()

__kernel void bitonic_start	(	__global T *restrict	vals,
		__global usize *restrict	keys )

Do the first half of the sort & merge stages.

Parameters

vals	Values
keys	Keys (permutations)

init()

__kernel void init	(	const __global T *restrict	in_vals,
		__global T *restrict	vals,
		__global usize *	keys,
		usize	n,
		usize	N )

Initializes the padded values and keys (permutations)

The permutations are initialized in such a way each point is permuted on itself

Parameters

in_vals	Unpadded and unsorted values
vals	Padded but unsorted values
keys	Permutations (keys)
n	Unpadded number of values.
N	Padded number of values.

inverse_keys()

__kernel void inverse_keys	(	const __global usize *restrict	keys,
		__global usize *	inv_keys,
		usize	n )

Compute the inversed permutations, which allows to know the original position of a key from the sorted one.

Parameters

keys	Direct permutations (from the unsorted position to the sorted one)
inv_keys	Inverse permutations (from the sorted position to the unsorted one)
n	Unpadded number of values.

SORT() [1/2]

SORT

(

__global

)

SORT() [2/2]

SORT

(

__local

)

SWAP() [1/2]

SWAP

(

__global

)

SWAP() [2/2]

SWAP

(

__local

)