KhronosGroup / KhronosGroup/OpenCL-Docs
reqd_work_group_size and non-uniform NDRanges
- Dominant language
- Python
- Stars
- 420
- Forks
- 131
- Avg merge
- 5d 13h
- Merged PRs (30d)
- 11
Description
The OpenCL C 2.0 specification currently reads (§6.7.2):
>The optional __attribute__((reqd_work_group_size(X, Y, Z))) is the work-group size that must be
used as the local_work_size argument to clEnqueueNDRangeKernel. This allows the compiler to
optimize the generated code appropriately for this kernel.
The only thing the compiler can know is that the work group size will be in the [{1,1,1},reqd_work_group_size] range when non-uniform NDRanges are in use which seems to defeat the purpose of having `reqd_work_group_size`.
FWIW, the non_uniform_work_group CTS tests do exercise the case where the attibute is specified but the NDRange is such that some work groups are smaller than the value specified with `reqd_work_group_size`.
Should we require NDRange uniformity when `reqd_work_group_size` is used instead?
Contributor guide
Assessment
This issue has not been assessed yet.