src/Utilities/CalculateNdRange.cpp

0001 // This file is part of the Acts project.
0002 //
0003 // Copyright (C) 2020 CERN for the benefit of the Acts project
0004 //
0005 // This Source Code Form is subject to the terms of the Mozilla Public
0006 // License, v. 2.0. If a copy of the MPL was not distributed with this
0007 // file, You can obtain one at http://mozilla.org/MPL/2.0/.
0008
0009 // SYCL plugin include(s)
0010 #include "Acts/Plugins/Sycl/Utilities/CalculateNdRange.hpp"
0011
0012 namespace Acts::Sycl {
0013 cl::sycl::nd_range<1> calculate1DimNDRange(const uint32_t numThreads,
0014                                            const uint32_t workGroupSize) {
0015   auto q = (numThreads + workGroupSize - 1) / workGroupSize;
0016   return cl::sycl::nd_range<1>{cl::sycl::range<1>(q * workGroupSize),
0017                                cl::sycl::range<1>(workGroupSize)};
0018 }
0019
0020 cl::sycl::nd_range<2> calculate2DimNDRange(const uint32_t numThreadsDim0,
0021                                            const uint32_t numThreadsDim1,
0022                                            const uint32_t workGroupSize) {
0023   uint32_t wgSizeDim0 = 1;
0024   uint32_t wgSizeDim1 = workGroupSize;
0025
0026   while (numThreadsDim1 < wgSizeDim1 && 1 < wgSizeDim1 && wgSizeDim1 % 2 == 0) {
0027     wgSizeDim1 /= 2;
0028     wgSizeDim0 *= 2;
0029   }
0030   auto q1 = (numThreadsDim0 + wgSizeDim0 + 1) / wgSizeDim0;
0031   auto q2 = (numThreadsDim1 + wgSizeDim1 + 1) / wgSizeDim1;
0032   return cl::sycl::nd_range<2>{
0033       cl::sycl::range<2>(q1 * wgSizeDim0, q2 * wgSizeDim1),
0034       cl::sycl::range<2>(wgSizeDim0, wgSizeDim1)};
0035 }
0036 }  // namespace Acts::Sycl