[SYCL][Doc] Remove overloads for kernel properties

This commit removes the parallel_for and single_task overloads from the
sycl_ext_oneapi_kernel_properties extension.

The sycl_ext_oneapi_enqueue_functions extension provides an alternative
interface for passing kernel properties that is more aligned with our
expected long-term direction: specifically, it separates run-time launch
properties from compile-time properties which affect kernel compilation,
and it introduces new concepts to avoid the need for extra arguments.

Signed-off-by: John Pennycook <[email protected]>

Author: Pennycook

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
@@ -232,141 +232,12 @@ SYCL implementations may introduce additional kernel properties. If any
 combinations of kernel attributes are invalid, this must be clearly documented
 as part of the new kernel property definition.
 
-=== 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 <typename KernelName, typename KernelType, typename PropertyList>
-  event single_task(PropertyList properties, const KernelType &kernelFunc);
-
-  template <typename KernelName, typename KernelType, typename PropertyList>
-  event single_task(event depEvent, PropertyList properties,
-                    const KernelType &kernelFunc);
-
-  template <typename KernelName, typename KernelType, typename PropertyList>
-  event single_task(const std::vector<event> &depEvents,
-                    PropertyList properties,
-                    const KernelType &kernelFunc);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(range<Dims> numWorkItems,
-                     Rest&&... rest);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(range<Dims> numWorkItems, event depEvent,
-                     PropertyList properties,
-                     Rest&&... rest);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(range<Dims> numWorkItems,
-                     const std::vector<event> &depEvents,
-                     PropertyList properties,
-                     Rest&&... rest);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(nd_range<Dims> executionRange,
-                     PropertyList properties,
-                     Rest&&... rest);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(nd_range<Dims> executionRange,
-                     event depEvent,
-                     PropertyList properties,
-                     Rest&&... rest);
-
-  template <typename KernelName, int Dims, typename PropertyList, typename... Rest>
-  event parallel_for(nd_range<Dims> executionRange,
-                     const std::vector<event> &depEvents,
-                     PropertyList properties,
-                     Rest&&... rest);
-}
-}
-
-namespace sycl {
-class handler {
- public:
-  template <typename KernelName, typename KernelType, typename PropertyList>
-  void single_task(PropertyList properties, const KernelType &kernelFunc);
-
-  template <typename KernelName, int dimensions, typename PropertyList, typename... Rest>
-  void parallel_for(range<dimensions> numWorkItems,
-                    PropertyList properties,
-                    Rest&&... rest);
-
-  template <typename KernelName, int dimensions, typename PropertyList, typename... Rest>
-  void parallel_for(nd_range<dimensions> executionRange,
-                    PropertyList properties,
-                    Rest&&... rest);
-
-  template <typename KernelName, typename WorkgroupFunctionType, int dimensions, typename PropertyList>
-  void parallel_for_work_group(range<dimensions> numWorkGroups,
-                               PropertyList properties,
-                               const WorkgroupFunctionType &kernelFunc);
-
-  template <typename KernelName, typename WorkgroupFunctionType, int dimensions, typename PropertyList>
-  void parallel_for_work_group(range<dimensions> numWorkGroups,
-                               range<dimensions> workGroupSize,
-                               PropertyList properties,
-                               const WorkgroupFunctionType &kernelFunc);
-}
-}
-```
-
-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
-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();
-```
-
-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:
-
-```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
 
-In other situations it may be useful to embed a kernel's properties directly
-into its type, to ensure that a kernel cannot be launched without a property
-that it depends upon for correctness.
+Compile-time kernel properties can be embedded directly into its type, to
+ensure that a kernel cannot be launched without a property that it depends upon
+for correctness.
 
 To enable this use-case, this extension adds a mechanism for implementations to
 extract a property list from a kernel functor, if a kernel functor declares
@@ -429,15 +300,6 @@ struct KernelFunctor {
 q.parallel_for(range<2>{16, 16}, KernelFunctor(a, b, c)).wait();
 ```
 
-If a kernel functor with embedded properties is enqueued for execution using an
-invocation function with a property list argument, the kernel is launched as-if
-the embedded properties and argument were combined. If the combined list
-contains any invalid combinations of properties, then this is an error: invalid
-combinations that can be detected at compile-time should be reported via a
-diagnostic; invalid combinations that can only be detected at run-time should
-result in an implementation throwing an `exception` with the `errc::invalid`
-error code.
-
 === Querying Properties in a Compiled Kernel
 
 Any properties embedded into a kernel type via a property list are reflected