KhronosGroup / KhronosGroup/OpenCL-Docs

reqd_work_group_size and non-uniform NDRanges

Open
#267 9 comments 0 reactions 1 assignee Claimed by @kpet View on GitHub
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

Open the contributing guide

Assessment

This issue has not been assessed yet.

Get new issues in your inbox

A short digest of beginner-friendly GitHub issues.