Skip to content

Commit 42277f4

Browse files
Aympabilluhad
Aympab
and
illuhad
authored
[KHR] Implement sycl_khr_max_work_group_queries extension (AdaptiveCpp#1736)
* Add max_num_work_group query for all backends * implement khr_max_num_work_groups * Add macro * revert accidental commit * add ocl backend * changed OCL backend to numeric limit * Impl device descriptor * change namespace name for extension * Revert "change namespace name for extension" This reverts commit ef288ae. * change namespace for extension * fix query namespace * merge changes * fix query namespace * Fix macro for extension * fix * fix #endif * update maximum size for OCL backend * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * fix namespace bracket * Add max_num_work_group query for all backends * implement khr_max_num_work_groups * Add macro * revert accidental commit * add ocl backend * changed OCL backend to numeric limit * Impl device descriptor * change namespace name for extension * Revert "change namespace name for extension" This reverts commit ef288ae. * change namespace for extension * fix query namespace * merge changes * fix query namespace * Fix macro for extension * fix * fix #endif * update maximum size for OCL backend * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * fix namespace bracket * rename extension to max_work_group_range * add max_work_group_range_size descriptor * rename query with queries * rename query with queries * Add max_num_work_group query for all backends * implement khr_max_num_work_groups * revert accidental commit * add ocl backend * changed OCL backend to numeric limit * Impl device descriptor * change namespace name for extension * Revert "change namespace name for extension" This reverts commit ef288ae. * change namespace for extension * fix query namespace * merge changes * fix query namespace * Fix macro for extension * fix * fix #endif * update maximum size for OCL backend * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * Change macro name for GET_INFO_KHR_EXTENSION Co-authored-by: Aksel Alpay <[email protected]> * fix namespace bracket * add ocl backend * Impl device descriptor * change namespace name for extension * Revert "change namespace name for extension" This reverts commit ef288ae. * change namespace for extension * merge changes * fix query namespace * Fix macro for extension * fix * fix #endif * fix namespace bracket * rename extension to max_work_group_range * add max_work_group_range_size descriptor --------- Co-authored-by: Aksel Alpay <[email protected]>
1 parent 02c42a2 commit 42277f4

File tree

11 files changed

+132
-1
lines changed

11 files changed

+132
-1
lines changed

‎include/hipSYCL/runtime/hardware.hpp‎

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -41,6 +41,10 @@ enum class device_support_aspect {
4141

4242
enum class device_uint_property {
4343
max_compute_units,
44+
max_work_group_range0,
45+
max_work_group_range1,
46+
max_work_group_range2,
47+
max_work_group_range_size,
4448
max_global_size0,
4549
max_global_size1,
4650
max_global_size2,

‎include/hipSYCL/sycl/device.hpp‎

Lines changed: 43 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -15,6 +15,7 @@
1515
#include <limits>
1616
#include <type_traits>
1717

18+
#include "info/device.hpp"
1819
#include "types.hpp"
1920
#include "aspect.hpp"
2021
#include "info/info.hpp"
@@ -305,7 +306,7 @@ HIPSYCL_SPECIALIZE_GET_INFO(device, device_type) {
305306
HIPSYCL_SPECIALIZE_GET_INFO(device, vendor_id)
306307
{
307308
return get_rt_device()->get_property(
308-
rt::device_uint_property::vendor_id);
309+
rt::device_uint_property::vendor_id);
309310
}
310311

311312
HIPSYCL_SPECIALIZE_GET_INFO(device, max_compute_units)
@@ -314,6 +315,47 @@ HIPSYCL_SPECIALIZE_GET_INFO(device, max_compute_units)
314315
rt::device_uint_property::max_compute_units);
315316
}
316317

318+
HIPSYCL_SPECIALIZE_GET_INFO_KHR_EXTENSION(device, max_work_group_range<1>)
319+
{
320+
std::size_t size0 = static_cast<std::size_t>(get_rt_device()->get_property(
321+
rt::device_uint_property::max_work_group_range0));
322+
return range<1>{size0};
323+
}
324+
325+
HIPSYCL_SPECIALIZE_GET_INFO_KHR_EXTENSION(device, max_work_group_range<2>)
326+
{
327+
std::size_t size0 = static_cast<std::size_t>(get_rt_device()->get_property(
328+
rt::device_uint_property::max_work_group_range0));
329+
std::size_t size1 = static_cast<std::size_t>(get_rt_device()->get_property(
330+
rt::device_uint_property::max_work_group_range1));
331+
if (get_rt_device()->get_property(
332+
rt::device_uint_property::needs_dimension_flip))
333+
return range<2>{size1, size0};
334+
else
335+
return range<2>{size0, size1};
336+
}
337+
338+
HIPSYCL_SPECIALIZE_GET_INFO_KHR_EXTENSION(device, max_work_group_range<3>)
339+
{
340+
std::size_t size0 = static_cast<std::size_t>(get_rt_device()->get_property(
341+
rt::device_uint_property::max_work_group_range0));
342+
std::size_t size1 = static_cast<std::size_t>(get_rt_device()->get_property(
343+
rt::device_uint_property::max_work_group_range1));
344+
std::size_t size2 = static_cast<std::size_t>(get_rt_device()->get_property(
345+
rt::device_uint_property::max_work_group_range2));
346+
if (get_rt_device()->get_property(
347+
rt::device_uint_property::needs_dimension_flip))
348+
return range<3>{size2, size1, size0};
349+
else
350+
return range<3>{size0, size1, size2};
351+
}
352+
353+
HIPSYCL_SPECIALIZE_GET_INFO_KHR_EXTENSION(device, max_work_group_range_size)
354+
{
355+
return get_rt_device()->get_property(
356+
rt::device_uint_property::max_work_group_range_size);
357+
}
358+
317359
HIPSYCL_SPECIALIZE_GET_INFO(device, max_work_item_dimensions)
318360
{ return 3; }
319361

‎include/hipSYCL/sycl/extensions.hpp‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -82,5 +82,6 @@
8282

8383
#define SYCL_KHR_DEFAULT_CONTEXT 1
8484
#define SYCL_KHR_QUEUE_EMPTY_QUERY 1
85+
#define SYCL_KHR_MAX_WORK_GROUP_QUERIES 1
8586

8687
#endif

‎include/hipSYCL/sycl/info/device.hpp‎

Lines changed: 11 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -175,6 +175,17 @@ namespace device {
175175
};
176176

177177
} // namespace info
178+
179+
namespace khr{
180+
namespace info{
181+
namespace device{
182+
template<int Dimensions = 3>
183+
struct max_work_group_range { using return_type = sycl::range<Dimensions>; };
184+
185+
struct max_work_group_range_size { using return_type = size_t; };
186+
}
187+
}
188+
}
178189
} // namespace sycl
179190
} // namespace hipsycl
180191

‎include/hipSYCL/sycl/info/info.hpp‎

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -27,4 +27,9 @@
2727
inline typename info::class_name::specialization::return_type \
2828
sycl::class_name::get_info<info::class_name::specialization>() const
2929

30+
#define HIPSYCL_SPECIALIZE_GET_INFO_KHR_EXTENSION(class_name, specialization) \
31+
template<> \
32+
inline typename khr::info::class_name::specialization::return_type \
33+
sycl::class_name::get_info<khr::info::class_name::specialization>() const
34+
3035
#endif

‎src/runtime/cuda/cuda_hardware_manager.cpp‎

Lines changed: 12 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -244,6 +244,18 @@ cuda_hardware_context::get_property(device_uint_property prop) const {
244244
case device_uint_property::max_compute_units:
245245
return _properties->multiProcessorCount;
246246
break;
247+
case device_uint_property::max_work_group_range0:
248+
return _properties->maxGridSize[0];
249+
break;
250+
case device_uint_property::max_work_group_range1:
251+
return _properties->maxGridSize[1];
252+
break;
253+
case device_uint_property::max_work_group_range2:
254+
return _properties->maxGridSize[2];
255+
break;
256+
case device_uint_property::max_work_group_range_size:
257+
return std::numeric_limits<std::size_t>::max();
258+
break;
247259
case device_uint_property::max_global_size0:
248260
return static_cast<std::size_t>(_properties->maxThreadsDim[0]) *
249261
_properties->maxGridSize[0];

‎src/runtime/hip/hip_hardware_manager.cpp‎

Lines changed: 12 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -281,6 +281,18 @@ hip_hardware_context::get_property(device_uint_property prop) const {
281281
case device_uint_property::max_compute_units:
282282
return _properties->multiProcessorCount;
283283
break;
284+
case device_uint_property::max_work_group_range0:
285+
return _properties->maxGridSize[0];
286+
break;
287+
case device_uint_property::max_work_group_range1:
288+
return _properties->maxGridSize[1];
289+
break;
290+
case device_uint_property::max_work_group_range2:
291+
return _properties->maxGridSize[2];
292+
break;
293+
case device_uint_property::max_work_group_range_size:
294+
return std::numeric_limits<std::size_t>::max();
295+
break;
284296
case device_uint_property::max_global_size0:
285297
return static_cast<std::size_t>(_properties->maxThreadsDim[0]) *
286298
_properties->maxGridSize[0];

‎src/runtime/ocl/ocl_hardware_manager.cpp‎

Lines changed: 16 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -20,6 +20,7 @@
2020
#include <CL/opencl.hpp>
2121
#include <cstddef>
2222
#include <array>
23+
// #include <limits>
2324
#include <optional>
2425

2526

@@ -290,6 +291,21 @@ std::size_t ocl_hardware_context::get_property(device_uint_property prop) const
290291
return static_cast<std::size_t>(
291292
info_query<CL_DEVICE_MAX_COMPUTE_UNITS, cl_uint>(_dev));
292293
break;
294+
case device_uint_property::max_work_group_range0:
295+
return static_cast<std::size_t>(
296+
std::numeric_limits<int>::max());
297+
break;
298+
case device_uint_property::max_work_group_range1:
299+
return static_cast<std::size_t>(
300+
std::numeric_limits<int>::max());
301+
break;
302+
case device_uint_property::max_work_group_range2:
303+
return static_cast<std::size_t>(
304+
std::numeric_limits<int>::max());
305+
break;
306+
case device_uint_property::max_work_group_range_size:
307+
return std::numeric_limits<std::size_t>::max();
308+
break;
293309
case device_uint_property::max_global_size0:
294310
return static_cast<std::size_t>(
295311
info_query<CL_DEVICE_MAX_WORK_ITEM_SIZES, std::vector<std::size_t>>(

‎src/runtime/omp/omp_hardware_manager.cpp‎

Lines changed: 12 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -136,6 +136,18 @@ omp_hardware_context::get_property(device_uint_property prop) const {
136136
// use this.
137137
return omp_get_num_procs();
138138
break;
139+
case device_uint_property::max_work_group_range0:
140+
return std::numeric_limits<std::size_t>::max();
141+
break;
142+
case device_uint_property::max_work_group_range1:
143+
return std::numeric_limits<std::size_t>::max();
144+
break;
145+
case device_uint_property::max_work_group_range2:
146+
return std::numeric_limits<std::size_t>::max();
147+
break;
148+
case device_uint_property::max_work_group_range_size:
149+
return std::numeric_limits<std::size_t>::max();
150+
break;
139151
case device_uint_property::max_global_size0:
140152
return std::numeric_limits<std::size_t>::max();
141153
break;

‎src/runtime/ze/ze_hardware_manager.cpp‎

Lines changed: 12 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -301,6 +301,18 @@ std::size_t ze_hardware_context::get_property(device_uint_property prop) const {
301301
case device_uint_property::max_compute_units:
302302
return _props.numSlices * _props.numSubslicesPerSlice * _props.numEUsPerSubslice;
303303
break;
304+
case device_uint_property::max_work_group_range0:
305+
return _compute_props.maxGroupSizeX;
306+
break;
307+
case device_uint_property::max_work_group_range1:
308+
return _compute_props.maxGroupSizeY;
309+
break;
310+
case device_uint_property::max_work_group_range2:
311+
return _compute_props.maxGroupSizeZ;
312+
break;
313+
case device_uint_property::max_work_group_range_size:
314+
return std::numeric_limits<std::size_t>::max();
315+
break;
304316
case device_uint_property::max_global_size0:
305317
return _compute_props.maxGroupSizeX * _compute_props.maxGroupCountX;
306318
break;

0 commit comments

Comments
 (0)