reduction_alg.hpp File Reference

#include "shambackends/DeviceBuffer.hpp"
#include "shambackends/sycl.hpp"
#include <memory>
#include <vector>

Include dependency graph for reduction_alg.hpp:

This graph shows which files directly or indirectly include this file:

Go to the source code of this file.

Classes
struct	reduc_ret_t< split_int >
	Return type of reduction algorithms. More...

Functions
template<class u_morton >
void	reduction_alg (sycl::queue &queue, u32 morton_count, std::unique_ptr< sycl::buffer< u_morton > > &buf_morton, u32 reduction_level, std::unique_ptr< sycl::buffer< u32 > > &buf_reduc_index_map, u32 &morton_leaf_count)
	Reduces a Morton tree on device.
template<class u_morton >
void	sycl_morton_remap_reduction (sycl::queue &queue, u32 morton_leaf_count, std::unique_ptr< sycl::buffer< u32 > > &buf_reduc_index_map, std::unique_ptr< sycl::buffer< u_morton > > &buf_morton, std::unique_ptr< sycl::buffer< u_morton > > &buf_leaf_morton)
	Remaps a Morton tree on device using a reduction index map.
template<class u_morton >
reduc_ret_t< u32 >	reduction_alg (const sham::DeviceScheduler_ptr &dev_sched, u32 morton_count, sham::DeviceBuffer< u_morton > &buf_morton, u32 reduction_level)
	Reduces a Morton tree on device.
template<class u_morton >
void	sycl_morton_remap_reduction (sham::DeviceQueue &queue, u32 morton_leaf_count, sham::DeviceBuffer< u32 > &buf_reduc_index_map, sham::DeviceBuffer< u_morton > &buf_morton, sham::DeviceBuffer< u_morton > &buf_leaf_morton)
	Remaps a Morton tree on device using a reduction index map.

Detailed Description

Author

Timothée David–Cléris (tim.s.nosp@m.hamr.nosp@m.ock@p.nosp@m.roto.nosp@m.n.me)

Definition in file reduction_alg.hpp.

Function Documentation

reduction_alg() [1/2]

template<class u_morton >

reduc_ret_t< u32 > reduction_alg	(	const sham::DeviceScheduler_ptr &	dev_sched,
		u32	morton_count,
		sham::DeviceBuffer< u_morton > &	buf_morton,
		u32	reduction_level
	)

Reduces a Morton tree on device.

This function will reduce a Morton tree represented by a device buffer buf_morton containing morton_count elements. The reduction is done using the following steps:

• generate a split table for the given tree
• perform a reduction on the tree using the split table
• compute the reduction index map
• remap the tree Morton codes using the reduction index map (using sycl_morton_remap_reduction)

The function takes a device scheduler dev_sched as input and will use it to schedule the kernel calls.

The function returns a struct of two elements:

• buf_reduc_index_map: a device buffer containing the reduction index map
• morton_leaf_count: the number of leaf nodes in the reduced tree

Parameters

dev_sched	a device scheduler to schedule the kernel calls
morton_count	the number of elements in the input Morton tree
buf_morton	a device buffer containing the Morton tree
reduction_level	the number of reduction levels to perform

Returns

a struct containing the reduction index map and the number of leaf nodes in the reduced tree

reduction_alg() [2/2]

template<class u_morton >

void reduction_alg	(	sycl::queue &	queue,
		u32	morton_count,
		std::unique_ptr< sycl::buffer< u_morton > > &	buf_morton,
		u32	reduction_level,
		std::unique_ptr< sycl::buffer< u32 > > &	buf_reduc_index_map,
		u32 &	morton_leaf_count
	)

Reduces a Morton tree on device.

This function will reduce a Morton tree represented by a device buffer buf_morton containing morton_count elements. The reduction is done using the following steps:

• generate a split table for the given tree
• perform a reduction on the tree using the split table
• compute the reduction index map
• remap the tree Morton codes using the reduction index map

The function takes a device scheduler dev_sched as input and will use it to schedule the kernel calls.

The function returns a struct of two elements:

• buf_reduc_index_map: a device buffer containing the reduction index map
• morton_leaf_count: the number of leaf nodes in the reduced tree

Parameters

queue	a SYCL queue to submit the kernels
morton_count	the number of elements in the input Morton tree
buf_morton	a device buffer containing the Morton tree
reduction_level	the number of reduction levels to perform
buf_reduc_index_map	a device buffer containing the reduction index map
morton_leaf_count	the number of leaf nodes in the reduced tree

sycl_morton_remap_reduction() [1/2]

template<class u_morton >

void sycl_morton_remap_reduction	(	sham::DeviceQueue &	queue,
		u32	morton_leaf_count,
		sham::DeviceBuffer< u32 > &	buf_reduc_index_map,
		sham::DeviceBuffer< u_morton > &	buf_morton,
		sham::DeviceBuffer< u_morton > &	buf_leaf_morton
	)

Remaps a Morton tree on device using a reduction index map.

This function takes a device buffer buf_reduc_index_map containing the reduction index map and a device buffer buf_morton containing the Morton tree. It will remap the Morton tree using the reduction index map and write the result to a device buffer buf_leaf_morton.

The function takes a device scheduler queue as input and will use it to schedule the kernel calls.

Parameters

queue	a device scheduler to schedule the kernel calls
morton_leaf_count	the number of leaf nodes in the reduced tree
buf_reduc_index_map	a device buffer containing the reduction index map
buf_morton	a device buffer containing the Morton tree
buf_leaf_morton	a device buffer that will contain the remapped Morton tree

Definition at line 565 of file reduction_alg.cpp.

Here is the call graph for this function:

sycl_morton_remap_reduction() [2/2]

template<class u_morton >

void sycl_morton_remap_reduction	(	sycl::queue &	queue,
		u32	morton_leaf_count,
		std::unique_ptr< sycl::buffer< u32 > > &	buf_reduc_index_map,
		std::unique_ptr< sycl::buffer< u_morton > > &	buf_morton,
		std::unique_ptr< sycl::buffer< u_morton > > &	buf_leaf_morton
	)

Remaps a Morton tree on device using a reduction index map.

This function takes a device buffer buf_reduc_index_map containing a reduction index map and a device buffer buf_morton containing a Morton tree and applies the reduction index map to the Morton tree. The output is stored in a new device buffer buf_leaf_morton containing the remapped Morton tree.

The function takes a device scheduler dev_sched as input and will use it to schedule the kernel calls.

Parameters

queue	a SYCL queue to submit the kernels
morton_leaf_count	the number of leaf nodes in the reduced tree
buf_reduc_index_map	a device buffer containing the reduction index map
buf_morton	a device buffer containing the Morton tree
buf_leaf_morton	a device buffer containing the remapped Morton tree