From f614f114f6888e5ff28edf286ca76ba6fd28f84c Mon Sep 17 00:00:00 2001 From: John Pennycook Date: Tue, 8 Oct 2024 15:54:07 +0100 Subject: [PATCH 1/3] [SYCL][Doc] Forbid multiple kernel property lists Updates the sycl_ext_oneapi_kernel_properties extension specification to align with both the current implementation and the behavior of the new sycl_ext_oneapi_enqueue_functions extension. Signed-off-by: John Pennycook --- ...sycl_ext_oneapi_kernel_properties.asciidoc | 24 ++++--------------- 1 file changed, 5 insertions(+), 19 deletions(-) diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc index d0ae2a0727046..0fe130588261c 100644 --- a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc +++ b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc @@ -21,7 +21,7 @@ == Notice [%hardbreaks] -Copyright (C) 2021 Intel Corporation. All rights reserved. +Copyright (C) 2024 Intel Corporation. All rights reserved. Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by @@ -37,7 +37,7 @@ https://github.com/intel/llvm/issues == Dependencies -This extension is written against the SYCL 2020 specification, Revision 4 and +This extension is written against the SYCL 2020 specification, Revision 9 and the following extensions: - link:sycl_ext_oneapi_properties.asciidoc[sycl_ext_oneapi_properties] @@ -387,12 +387,6 @@ class handler { } ``` -Passing a property list as an argument in this way allows properties to be -associated with a kernel function without modifying its type. This enables -the same kernel function (e.g. a lambda) to be submitted multiple times with -different properties, or for libraries building on SYCL to add properties -(e.g. for performance reasons) to user-provided kernel functions. - All the properties defined in this extension have compile-time values. However, an implementation may support additional properties which could have run-time values. When this occurs, the `properties` parameter may be a property list @@ -412,18 +406,10 @@ q.parallel_for(range<2>{16, 16}, properties, [=](id<2> i) { }).wait(); ``` -NOTE: It is currently not possible to use the same kernel function in two -commands with different properties. For example, the following will result in an -error at compile-time: +A kernel function can only be associated with one set of kernel properties. +If a kernel function is associated with more than one set of kernel properties, +the implementation must issue a diagnostic. -```c++ - auto kernelFunc = [=](){}; - q.single_task(kernelFunc); - q.single_task( - sycl::ext::oneapi::experimental::properties{ - sycl::ext::oneapi::experimental::sub_group_size<8>}, - kernelFunc); -``` == Embedding Properties into a Kernel From 925c7d297333ac346eaba924ff25f4582efdfbfb Mon Sep 17 00:00:00 2001 From: John Pennycook Date: Tue, 8 Oct 2024 16:03:47 +0100 Subject: [PATCH 2/3] [SYCL][Doc] Remove kernel property overloads The enqueue functions in sycl_ext_oneapi_enqueue_functions provide an alternative interface for attaching kernel properties to a launch configuration, removing the need for these overloads. Signed-off-by: John Pennycook --- ...sycl_ext_oneapi_kernel_properties.asciidoc | 116 ------------------ 1 file changed, 116 deletions(-) diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc index 0fe130588261c..1680d47461a7a 100644 --- a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc +++ b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc @@ -295,122 +295,6 @@ by the property, the implementation must throw a synchronous exception with the |=== -=== Adding a Property List to a Kernel Launch - -To enable properties to be associated with kernels, this extension adds -new overloads to each of the variants of `single_task`, `parallel_for` and -`parallel_for_work_group` defined in the `queue` and `handler` classes. These -new overloads accept a `sycl::ext::oneapi::experimental::properties` argument. For -variants accepting a parameter pack, the `sycl::ext::oneapi::experimental::properties` -argument is inserted immediately prior to the parameter pack; for variants not -accepting a parameter pack, the `sycl::ext::oneapi::experimental::properties` argument is -inserted immediately prior to the kernel function. - -The overloads introduced by this extension are listed below: -```c++ -namespace sycl { -class queue { - public: - template - event single_task(PropertyList properties, const KernelType &kernelFunc); - - template - event single_task(event depEvent, PropertyList properties, - const KernelType &kernelFunc); - - template - event single_task(const std::vector &depEvents, - PropertyList properties, - const KernelType &kernelFunc); - - template - event parallel_for(range numWorkItems, - Rest&&... rest); - - template - event parallel_for(range numWorkItems, event depEvent, - PropertyList properties, - Rest&&... rest); - - template - event parallel_for(range numWorkItems, - const std::vector &depEvents, - PropertyList properties, - Rest&&... rest); - - template - event parallel_for(nd_range executionRange, - PropertyList properties, - Rest&&... rest); - - template - event parallel_for(nd_range executionRange, - event depEvent, - PropertyList properties, - Rest&&... rest); - - template - event parallel_for(nd_range executionRange, - const std::vector &depEvents, - PropertyList properties, - Rest&&... rest); -} -} - -namespace sycl { -class handler { - public: - template - void single_task(PropertyList properties, const KernelType &kernelFunc); - - template - void parallel_for(range numWorkItems, - PropertyList properties, - Rest&&... rest); - - template - void parallel_for(nd_range executionRange, - PropertyList properties, - Rest&&... rest); - - template - void parallel_for_work_group(range numWorkGroups, - PropertyList properties, - const WorkgroupFunctionType &kernelFunc); - - template - void parallel_for_work_group(range numWorkGroups, - range workGroupSize, - PropertyList properties, - const WorkgroupFunctionType &kernelFunc); -} -} -``` - -All the properties defined in this extension have compile-time values. However, -an implementation may support additional properties which could have run-time -values. When this occurs, the `properties` parameter may be a property list -containing a mix of both run-time and compile-time values, and a SYCL -implementation should respect both run-time and compile-time information when -determining the correct way to launch a kernel. However, only compile-time -information can modify the compilation of the kernel function itself. - -A simple example of using this extension to set a required work-group size -and required sub-group size is given below: - -```c++ -sycl::ext::oneapi::experimental::properties properties{sycl::ext::oneapi::experimental::work_group_size<8, 8>, - sycl::ext::oneapi::experimental::sub_group_size<8>}; -q.parallel_for(range<2>{16, 16}, properties, [=](id<2> i) { - a[i] = b[i] + c[i]; -}).wait(); -``` - -A kernel function can only be associated with one set of kernel properties. -If a kernel function is associated with more than one set of kernel properties, -the implementation must issue a diagnostic. - - == Embedding Properties into a Kernel In other situations it may be useful to embed a kernel's properties directly From 974aec94af2ab81014895cf961895b5d2c06fc29 Mon Sep 17 00:00:00 2001 From: John Pennycook Date: Wed, 9 Oct 2024 08:06:25 +0100 Subject: [PATCH 3/3] Fix copyright year (again) Signed-off-by: John Pennycook --- .../experimental/sycl_ext_oneapi_kernel_properties.asciidoc | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc index 1680d47461a7a..e88279d6c97dd 100644 --- a/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc +++ b/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_properties.asciidoc @@ -21,7 +21,7 @@ == Notice [%hardbreaks] -Copyright (C) 2024 Intel Corporation. All rights reserved. +Copyright (C) 2021 Intel Corporation. All rights reserved. Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by