cory-master/BabelStream
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

[SYCL 2020] Remove work-group heuristic for reduction as unused

Browse Source
This commit is contained in:
Tom Deakin 2021-02-08 11:07:38 +00:00
parent fb7ef884c6
commit 1517101ceb
2 changed files with 0 additions and 17 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

13
SYCLStream.cpp
View File

@ -27,22 +27,9 @@ SYCLStream<T>::SYCLStream(const int ARRAY_SIZE, const int device_index)
sycl::device dev = devices[device_index]; sycl::device dev = devices[device_index];
// Determine sensible dot kernel NDRange configuration
if (dev.is_cpu())
{
dot_num_groups = dev.get_info<sycl::info::device::max_compute_units>();
dot_wgsize = dev.get_info<sycl::info::device::native_vector_width_double>() * 2;
}
else
{
dot_num_groups = dev.get_info<sycl::info::device::max_compute_units>() * 4;
dot_wgsize = dev.get_info<sycl::info::device::max_work_group_size>();
}
// Print out device information // Print out device information
std::cout << "Using SYCL device " << getDeviceName(device_index) << std::endl; std::cout << "Using SYCL device " << getDeviceName(device_index) << std::endl;
std::cout << "Driver: " << getDeviceDriver(device_index) << std::endl; std::cout << "Driver: " << getDeviceDriver(device_index) << std::endl;
std::cout << "Reduction kernel config: " << dot_num_groups << " groups of size " << dot_wgsize << std::endl;
queue = new sycl::queue(dev, sycl::async_handler{[&](sycl::exception_list l) queue = new sycl::queue(dev, sycl::async_handler{[&](sycl::exception_list l)
{ {

4
SYCLStream.h
View File

@ -29,10 +29,6 @@ class SYCLStream : public Stream<T>
sycl::buffer<T> *d_c; sycl::buffer<T> *d_c;
sycl::buffer<T> *d_sum; sycl::buffer<T> *d_sum;
// NDRange configuration for the dot kernel
size_t dot_num_groups;
size_t dot_wgsize;
public: public:
SYCLStream(const int, const int); SYCLStream(const int, const int);