[SYCL][Doc] Add named sub-group sizes extension #5714
Conversation
This extension aims to simplify the process of using sub-groups by introducing the notion of named sub-group sizes, allowing developers to request a sub-group size that meets certain requirements at host compile-time and deferring the selection of a specific sub-group size until the kernel is compiled for a specific device. Signed-off-by: John Pennycook <[email protected]>
Pennycook
added
the
spec extension
Misunderstood the directory structure. This shouldn't be in "experimental" yet, because it isn't implemented.
sycl/doc/extensions/proposed/sycl_ext_oneapi_named_sub_group_sizes.asciidoc
Outdated
Show resolved
Hide resolved
| enum named_sub_group_size : uint32_t { | ||
| primary = /* implementation-defined */, | ||
| automatic = /* implementation-defined */, | ||
| }; |
There was a problem hiding this comment.
Because this is an unscoped enum, the names primary and automatic are also in the sycl::ext::oneapi::experimental namespace. I suspect that was not your intent? Unfortunately, you cannot just make this a scoped enum because then the type of primary and automatic are no longer implicitly convertible to uint32_t. Maybe this is better?
static constexpr uint32_t named_sub_group_primary = /*unspecified*/;
static constexpr uint32_t named_sub_group_automatic = /*unspecified*/;
There was a problem hiding this comment.
Ah, you're right... I avoided making it a scoped enum precisely because of the uint32_t issue.
What about:
struct named_sub_group_size {
static constexpr uint32_t primary = /* unspecified */;
static constexpr uint32_t automatic = /* unspecified */;
};I don't have a strong preference here, just floating this idea because I think it would allow the same syntax (named_sub_group_size::primary).
There was a problem hiding this comment.
Yes, that looks good too. I don't have a strong preference between the two styles.
There was a problem hiding this comment.
I've gone with the struct approach in 23cca73, simply because it meant changing fewer lines.
I'd be happy to revisit this later.
| units together. | ||
|
|
||
|
|
||
| === Compiler flags |
There was a problem hiding this comment.
This should not be a normative section of the specification. Other implementations should be allowed to choose their own command line switches. Maybe retitle the section "DPC++ compiler flags" and start with these sentences:
This non-normative section describes command line flags that the DPC++ compiler supports. Other compilers are free to provide their own command line flags (if any).
| specification are implemented in {dpcpp}, but they are not finalized and may | ||
| change incompatibly in future versions of {dpcpp} without prior notice. | ||
| *Shipping software products should not rely on APIs defined in this | ||
| specification.* |
There was a problem hiding this comment.
This should be updated to the "proposed" language from the template.
sycl/doc/extensions/proposed/sycl_ext_oneapi_named_sub_group_sizes.asciidoc
Outdated
Show resolved
Hide resolved
Co-authored-by: Greg Lueck <[email protected]>
Co-authored-by: Greg Lueck <[email protected]>
This extension aims to simplify the process of using sub-groups by
introducing the notion of named sub-group sizes, allowing developers
to request a sub-group size that meets certain requirements at host
compile-time and deferring the selection of a specific sub-group size
until the kernel is compiled for a specific device.
Signed-off-by: John Pennycook [email protected]