Merge from 'sycl' to 'sycl-web'

Author: iclsrc

clang/lib/Driver/ToolChains/Clang.cpp

@@ -5257,7 +5257,7 @@ void Clang::ConstructJob(Compilation &C, const JobAction &JA,
                       options::OPT_fno_sycl_early_optimizations,
                       !IsFPGASYCLOffloadDevice))
       CmdArgs.push_back("-fno-sycl-early-optimizations");
-    else if (RawTriple.isSPIR() || IsSYCLNativeCPU) {
+    else if (RawTriple.isSPIR()) {
       // Set `sycl-opt` option to configure LLVM passes for SPIR target
       CmdArgs.push_back("-mllvm");
       CmdArgs.push_back("-sycl-opt");

clang/test/Driver/sycl-native-cpu.cpp

@@ -17,3 +17,8 @@
 // CHECK-WIN: {{.*}}"-fsycl-is-device"{{.*}}"-gcodeview"
 // CHECK-WIN-DAG: {{.*}}"-fsycl-is-host"{{.*}}"-gcodeview"
 // CHECK-WIN-NOT: dwarf
+
+// checks that -sycl-opt is not enabled by default on NativeCPU so that the full llvm optimization is enabled
+// RUN:   %clang -fsycl -fsycl-targets=native_cpu -### %s 2>&1 | FileCheck -check-prefix=CHECK-OPTS %s
+// CHECK-OPTS-NOT: -sycl-opt
+

sycl/doc/EnvironmentVariables.md

@@ -9,7 +9,7 @@ compiler and runtime.
 | -------------------- | ------ | ----------- |
 | `ONEAPI_DEVICE_SELECTOR` | [See below.](#oneapi_device_selector)  | This device selection environment variable can be used to limit the choice of devices available when the SYCL-using application is run.  Useful for limiting devices to a certain type (like GPUs or accelerators) or backends (like Level Zero or OpenCL).  This device selection mechanism is replacing `SYCL_DEVICE_FILTER` .  The `ONEAPI_DEVICE_SELECTOR` syntax is shared with OpenMP and also allows sub-devices to be chosen. [See below.](#oneapi_device_selector) for a full description.   |
 | `SYCL_DEVICE_FILTER` (deprecated) | `backend:device_type:device_num` | Please use `ONEAPI_DEVICE_SELECTOR` environment variable instead. See section [`SYCL_DEVICE_FILTER`](#sycl_device_filter) below for `SYCL_DEVICE_FILTER` description. |
-| `SYCL_DEVICE_ALLOWLIST` | See [below](#sycl_device_allowlist) | Filter out devices that do not match the pattern specified. `BackendName` accepts `host`, `opencl`, `level_zero` or `cuda`. `DeviceType` accepts `host`, `cpu`, `gpu` or `acc`. `DeviceVendorId` accepts uint32_t in hex form (`0xXYZW`). `DriverVersion`, `PlatformVersion`, `DeviceName` and `PlatformName` accept regular expression. Special characters, such as parenthesis, must be escaped. DPC++ runtime will select only those devices which satisfy provided values above and regex. More than one device can be specified using the piping symbol "\|".|
+| `SYCL_DEVICE_ALLOWLIST` | See [below](#sycl_device_allowlist) | Filter out devices that do not match the pattern specified. `BackendName` accepts `host`, `opencl`, `level_zero`, `native_cpu` or `cuda`. `DeviceType` accepts `host`, `cpu`, `gpu` or `acc`. `DeviceVendorId` accepts uint32_t in hex form (`0xXYZW`). `DriverVersion`, `PlatformVersion`, `DeviceName` and `PlatformName` accept regular expression. Special characters, such as parenthesis, must be escaped. DPC++ runtime will select only those devices which satisfy provided values above and regex. More than one device can be specified using the piping symbol "\|".|
 | `SYCL_DISABLE_PARALLEL_FOR_RANGE_ROUNDING` | Any(\*) | Disables automatic rounding-up of `parallel_for` invocation ranges. |
 | `SYCL_CACHE_DIR` | Path | Path to persistent cache root directory. Default values are `%AppData%\libsycl_cache` for Windows and `$XDG_CACHE_HOME/libsycl_cache` on Linux, if `XDG_CACHE_HOME` is not set then `$HOME/.cache/libsycl_cache`. When none of the environment variables are set SYCL persistent cache is disabled. |
 | `SYCL_CACHE_DISABLE_PERSISTENT (deprecated)` | Any(\*) | Has no effect. |
@@ -42,7 +42,7 @@ ONEAPI_DEVICE_SELECTOR = <selector-string>
 <accept-filter> ::= <term>
 <discard-filter> ::= !<term>
 <term> ::= <backend>:<devices>
-<backend> ::= { * | level_zero | opencl | cuda | hip }  // case insensitive
+<backend> ::= { * | level_zero | opencl | cuda | hip | native_cpu }  // case insensitive
 <devices> ::= <device>[,<device>...]
 <device> ::= { * | cpu | gpu | fpga | <num> | <num>.<num> | <num>.* | *.* | <num>.<num>.<num> | <num>.<num>.* | <num>.*.* | *.*.*  }  // case insensitive
 ```
@@ -58,13 +58,13 @@ The device indices are zero-based and are unique only within a backend. Therefor
 
 Additionally, if a sub-device is chosen (via numeric index or wildcard), then an additional layer of partitioning can be specified. In other words, a sub-sub-device can be selected. Like sub-devices, this is done with a period ( `.` ) and a sub-sub-device specifier which is a wildcard symbol ( `*` ) or a numeric index.  Example `ONEAPI_DEVICE_SELECTOR=level_zero:0.*.*` would partition device 0 into sub-devices and then partition each of those into sub-sub-devices. The range of grandchild sub-sub-devices would be the final devices available to the app, neither device 0, nor its child partitions would be in that list.
 
-Lastly, a filter in the grammar can be thought of as a term in conjuction with an action that is taken on all devices that are selected by the term. The action can be an accept action or a discard action. Based on the action, a filter can be an accept filter or a discard filter.
+Lastly, a filter in the grammar can be thought of as a term in conjunction with an action that is taken on all devices that are selected by the term. The action can be an accept action or a discard action. Based on the action, a filter can be an accept filter or a discard filter.
 The string `<term>` represents an accept filter and the string `!<term>` represents a discard filter. The underlying term is the same but they perform different actions on the matching devices list.
 For example, `!opencl:*` discards all devices of the opencl backend from the list of available devices. The discarding filters, if there are any, must all appear at the end of the selector string.
 When one or more filters accept a device and one or more filters discard the device, the latter have priority and the device is ultimately not made available to the user. This allows the user to provide selector strings such as `*:gpu;!cuda:*` that accepts all gpu devices except those with a CUDA backend.
 Furthermore, if the value of this environment variable only has discarding filters, an accepting filter that matches all devices, but not sub-devices and sub-sub-devices, will be implicitly included in the
 environment variable to allow the user to specify only the list of devices that must not be made available. Therefore, `!*:cpu` will accept all devices except those that are of the cpu type and `opencl:*;!*:cpu`
-will accept all devices of the opencl backend exept those that are of the opencl backend and of the cpu type. It is legal to have a rejection filter even if it specifies devices have already been omitted by previous filters in the selection string. Doing so has no effect; the rejected devices are still omitted.
+will accept all devices of the opencl backend except those that are of the opencl backend and of the cpu type. It is legal to have a rejection filter even if it specifies devices have already been omitted by previous filters in the selection string. Doing so has no effect; the rejected devices are still omitted.
 
 The following examples further illustrate the usage of this environment variable:
 
@@ -290,7 +290,7 @@ variables in production code.</span>
 | `SYCL_PI_LEVEL_ZERO_EXPOSE_CSLICE_IN_AFFINITY_PARTITIONING` (Deprecated) | Integer | When set to non-zero value exposes compute slices as sub-sub-devices in `sycl::info::partition_property::partition_by_affinity_domain` partitioning scheme. Default is zero meaning that they are only exposed when partitioning by `sycl::info::partition_property::ext_intel_partition_by_cslice`. This option is introduced for compatibility reasons and is immediately deprecated. New code must not rely on this behavior. Also note that even if sub-sub-device was created using `partition_by_affinity_domain` it would still be reported as created via partitioning by compute slices. |
 | `SYCL_PI_LEVEL_ZERO_COMMANDLISTS_CLEANUP_THRESHOLD` | Integer | If non-negative then the threshold is set to this value. If negative, the threshold is set to INT_MAX. Whenever the number of command lists in a queue exceeds this threshold, an attempt is made to cleanup completed command lists for their subsequent reuse. The default is 20. |
 | `SYCL_PI_LEVEL_ZERO_IMMEDIATE_COMMANDLISTS_EVENT_CLEANUP_THRESHOLD` | Integer | If non-negative then the threshold is set to this value. If negative, the threshold is set to INT_MAX. Whenever the number of events associated with an immediate command list exceeds this threshold, a check is made for signaled events and these events are recycled. Setting this threshold low causes events to be checked more often, which could result in unneeded events being recycled sooner. However, more frequent event status checks may cost time. The default is 1000. |
-| `SYCL_PI_LEVEL_ZERO_USM_RESIDENT` | Integer | Bit-mask controls if/where to make USM allocations resident at the time of allocation. Input value is of the form 0xHSD, where 4-bits of D control device allocations, 4-bits of S control shared allocations, and 4-bits of H control host allocations. Each 4-bit componenet is holding one of the following values: "0" - then no special residency is forced, "1" - then allocation is made resident at the device of allocation, or "2" - then allocation is made resident on all devices in the context of allocation that have P2P access to the device of allocation. Default is 0x002, i.e. force full residency for device allocations only. |
+| `SYCL_PI_LEVEL_ZERO_USM_RESIDENT` | Integer | Bit-mask controls if/where to make USM allocations resident at the time of allocation. Input value is of the form 0xHSD, where 4-bits of D control device allocations, 4-bits of S control shared allocations, and 4-bits of H control host allocations. Each 4-bit component is holding one of the following values: "0" - then no special residency is forced, "1" - then allocation is made resident at the device of allocation, or "2" - then allocation is made resident on all devices in the context of allocation that have P2P access to the device of allocation. Default is 0x002, i.e. force full residency for device allocations only. |
 | `SYCL_PI_LEVEL_ZERO_USE_NATIVE_USM_MEMCPY2D` | Integer | When set to a positive value enables the use of Level Zero USM 2D memory copy operations. Default is 0. |
 
 ## Debugging variables for CUDA Plugin

sycl/doc/extensions/experimental/sycl_ext_matrix/sycl_ext_oneapi_matrix.asciidoc

@@ -915,9 +915,11 @@ architecture::intel_gpu_dg2_g11, architecture::intel_gpu_dg2_g12`
 `matrix_type::fp32`   .2+|  +<=+ 8 |  16   .2+|  16 |
 `architecture::intel_gpu_pvc`|8| `architecture::intel_gpu_dg2_g10,
 architecture::intel_gpu_dg2_g11, architecture::intel_gpu_dg2_g12`
-.2+|  `matrix_type::bf16`       .2+|  `matrix_type::bf16`   .2+|
-`matrix_type::fp32`   .2+|  +<=+ 8 |  16   .2+|  16 |
-`architecture::intel_gpu_pvc` |8| `architecture::intel_gpu_dg2_g10,
+.4+|  `matrix_type::bf16`       .4+|  `matrix_type::bf16`   .4+|
+`matrix_type::fp32`  | 16 | 16 | 16 .3+|`architecture::intel_gpu_pvc` |
+32 | 64 | 16
+.2+|  +<=+ 8 |  16   .2+|  16
+|8 | `architecture::intel_gpu_dg2_g10,
 architecture::intel_gpu_dg2_g11, architecture::intel_gpu_dg2_g12`
 |  `matrix_type::tf32`       |  `matrix_type::tf32`   |
 `matrix_type::fp32`   |  +<=+ 8 |  16   |  8 |

sycl/include/sycl/accessor.hpp

@@ -982,9 +982,9 @@ class image_accessor
 
     device Device = getDeviceFromHandler(CommandGroupHandlerRef);
     if (!Device.has(aspect::ext_intel_legacy_image))
-      throw feature_not_supported(
-          "SYCL 1.2.1 images are not supported by this device.",
-          PI_ERROR_INVALID_OPERATION);
+      throw sycl::exception(
+          sycl::errc::feature_not_supported,
+          "SYCL 1.2.1 images are not supported by this device.");
   }
 #endif
 

sycl/include/sycl/ext/oneapi/group_local_memory.hpp

@@ -11,7 +11,7 @@
 #include <sycl/detail/defines_elementary.hpp> // for __SYCL_ALWAYS_INLINE
 #include <sycl/detail/pi.h>                   // for PI_ERROR_INVALID_OPERA...
 #include <sycl/detail/type_traits.hpp>        // for is_group
-#include <sycl/exception.hpp>                 // for feature_not_supported
+#include <sycl/exception.hpp>                 // for exception
 #include <sycl/ext/intel/usm_pointers.hpp>    // for multi_ptr
 
 #include <type_traits> // for enable_if_t
@@ -42,9 +42,9 @@ std::enable_if_t<
   }
   return reinterpret_cast<__attribute__((opencl_local)) T *>(AllocatedMem);
 #else
-  throw feature_not_supported(
-      "sycl_ext_oneapi_local_memory extension is not supported on host device",
-      PI_ERROR_INVALID_OPERATION);
+  throw sycl::exception(
+      sycl::errc::feature_not_supported,
+      "sycl_ext_oneapi_local_memory extension is not supported on host");
 #endif
 }
 
@@ -64,9 +64,9 @@ std::enable_if_t<
   // Silence unused variable warning
   (void)g;
   [&args...] {}();
-  throw feature_not_supported(
-      "sycl_ext_oneapi_local_memory extension is not supported on host device",
-      PI_ERROR_INVALID_OPERATION);
+  throw sycl::exception(
+      sycl::errc::feature_not_supported,
+      "sycl_ext_oneapi_local_memory extension is not supported on host");
 #endif
 }
 } // namespace ext::oneapi

sycl/include/sycl/handler.hpp

@@ -1261,9 +1261,10 @@ class __SYCL_EXPORT handler {
                   "first argument of sycl::item type, or of a type which is "
                   "implicitly convertible from sycl::item");
 
+    using RefLambdaArgType = std::add_lvalue_reference_t<LambdaArgType>;
     static_assert(
-        (std::is_invocable_v<KernelType, LambdaArgType> ||
-         std::is_invocable_v<KernelType, LambdaArgType, kernel_handler>),
+        (std::is_invocable_v<KernelType, RefLambdaArgType> ||
+         std::is_invocable_v<KernelType, RefLambdaArgType, kernel_handler>),
         "SYCL kernel lambda/functor has an unexpected signature, it should be "
         "invocable with sycl::item and optionally sycl::kernel_handler");
 #endif

sycl/source/detail/device_info.hpp

@@ -763,6 +763,10 @@ struct get_device_info_impl<
            matrix_type::fp32, matrix_type::fp32},
           {8, 0, 0, 0, 16, 16, matrix_type::bf16, matrix_type::bf16,
            matrix_type::fp32, matrix_type::fp32},
+          {0, 0, 0, 16, 16, 16, matrix_type::bf16, matrix_type::bf16,
+           matrix_type::fp32, matrix_type::fp32},
+          {0, 0, 0, 32, 64, 16, matrix_type::bf16, matrix_type::bf16,
+           matrix_type::fp32, matrix_type::fp32},
           {8, 0, 0, 0, 16, 8, matrix_type::tf32, matrix_type::tf32,
            matrix_type::fp32, matrix_type::fp32},
       };

sycl/source/detail/event_impl.cpp

@@ -464,11 +464,13 @@ void event_impl::setSubmissionTime() {
     if (QueueImplPtr Queue = MQueue.lock()) {
       try {
         MSubmitTime = Queue->getDeviceImplPtr()->getCurrentDeviceTime();
-      } catch (feature_not_supported &e) {
-        throw sycl::exception(
-            make_error_code(errc::profiling),
-            std::string("Unable to get command group submission time: ") +
-                e.what());
+      } catch (sycl::exception &e) {
+        if (e.code() == sycl::errc::feature_not_supported)
+          throw sycl::exception(
+              make_error_code(errc::profiling),
+              std::string("Unable to get command group submission time: ") +
+                  e.what());
+        std::rethrow_exception(std::current_exception());
       }
     }
   } else {

sycl/source/detail/program_impl.cpp

@@ -36,10 +36,10 @@ program_impl::program_impl(ContextImplPtr Context,
                            const property_list &PropList)
     : MContext(Context), MDevices(DeviceList), MPropList(PropList) {
   if (Context->getDevices().size() > 1) {
-    throw feature_not_supported(
+    throw sycl::exception(
+        sycl::errc::feature_not_supported,
         "multiple devices within a context are not supported with "
-        "sycl::program and sycl::kernel",
-        PI_ERROR_INVALID_OPERATION);
+        "sycl::program and sycl::kernel");
   }
 }
 
@@ -65,10 +65,10 @@ program_impl::program_impl(
 
   MContext = ProgramList[0]->MContext;
   if (MContext->getDevices().size() > 1) {
-    throw feature_not_supported(
+    throw sycl::exception(
+        sycl::errc::feature_not_supported,
         "multiple devices within a context are not supported with "
-        "sycl::program and sycl::kernel",
-        PI_ERROR_INVALID_OPERATION);
+        "sycl::program and sycl::kernel");
   }
   MDevices = ProgramList[0]->MDevices;
   std::vector<device> DevicesSorted;