Merge branch 'sycl' into mark-split-exceptions

Author: ayylol

.github/workflows/scorecard.yml

@@ -57,6 +57,6 @@ jobs:
 
       # Upload the results to GitHub's code scanning dashboard.
       - name: "Upload to code-scanning"
-        uses: github/codeql-action/upload-sarif@e2b3eafc8d227b0241d48be5f425d47c2d750a13 # v3.26.10
+        uses: github/codeql-action/upload-sarif@662472033e021d55d94146f66f6058822b0b39fd # v3.27.0
         with:
           sarif_file: results.sarif

.github/workflows/sycl-nightly.yml

@@ -198,7 +198,7 @@ jobs:
           echo "TAG=$(date +'%Y-%m-%d')-${GITHUB_SHA::7}" >> "$GITHUB_OUTPUT"
         fi
     - name: Upload binaries
-      uses: softprops/action-gh-release@c062e08bd532815e2082a85e87e3ef29c3e6d191
+      uses: softprops/action-gh-release@e7a8f85e1c67a31e6ed99a94b41bd0b71bbee6b8
       with:
         files: |
           sycl_linux.tar.gz

devops/cts_exclude_filter_L0_GPU

@@ -3,5 +3,3 @@ kernel_bundle
 marray
 # fix: https://github.com/KhronosGroup/SYCL-CTS/pull/964
 accessor_legacy
-# CMPLRLLVM-62822
-multi_ptr

devops/cts_exclude_filter_OCL_CPU

@@ -7,5 +7,3 @@ math_builtin_api
 hierarchical
 # fix: https://github.com/KhronosGroup/SYCL-CTS/pull/964
 accessor_legacy
-# CMPLRLLVM-62822
-multi_ptr

llvm/docs/requirements-hashed.txt

@@ -1,4 +1,4 @@
-sphinx==8.0.2
+sphinx==8.1.3
 docutils==0.21.2
 sphinx-markdown-tables==0.0.17
 recommonmark==0.7.1

llvm/docs/requirements.txt

@@ -1,7 +1,7 @@
-# commit 3d58884b4939d9bd095c917f8dd823ac8486684c
-# Merge: 6ade245e b0bd146a
+# commit fa8cc8ec16c1a2cf0926cc64026edc6a254ff0c2
+# Merge: 3d58884b 1984ceb1
 # Author: aarongreig <[email protected]>
-# Date:   Thu Oct 31 14:05:55 2024 +0000
-#     Merge pull request #2228 from nrspruit/copy_engine_refactor
-#     [L0] Refactor Copy Engine Usage checks for Performance
-set(UNIFIED_RUNTIME_TAG 3d58884b4939d9bd095c917f8dd823ac8486684c)
+# Date:   Fri Nov 1 10:03:37 2024 +0000
+#     Merge pull request #1833 from ayylol/old-opencl-check
+#     [OpenCL] Add OpenCL version check for independent forward progress query
+set(UNIFIED_RUNTIME_TAG fa8cc8ec16c1a2cf0926cc64026edc6a254ff0c2)

sycl/cmake/modules/UnifiedRuntimeTag.cmake

@@ -10,26 +10,20 @@
 //  1. Add a new enumerator to
 //     `sycl::ext::oneapi::experimental::detail::PropKind` representing the new
 //     property. Increment
-//     `sycl::ext::oneapi::experimental::detail::PropKind::PropKindSize`
-//  2. Define property key class with `value_t` that must be `property_value`
-//     with the first template argument being the property class itself. The
-//     name of the key class must be the property name suffixed by `_key`, i.e.
-//     for a property `foo` the class should be named `foo_key`.
+//     `sycl::ext::oneapi::experimental::detail::PropKind::PropKindSize`.
+//  2. Define property key class inherited from
+//     `detail::compile_time_property_key` with `value_t` that must be
+//     `property_value` with the first template argument being the property
+//     class itself. The name of the key class must be the property name
+//     suffixed by `_key`, i.e. for a property `foo` the class should be named
+//     `foo_key`.
 //  3. Add an `inline constexpr` variable in the same namespace as the property
 //     key. The variable should have the same type as `value_t` of the property
 //     class, e.g. for a property `foo`, there should be a definition
 //     `inline constexpr foo_key::value_t foo`.
-//  4. Specialize `sycl::ext::oneapi::experimental::is_property_key` and
-//     `sycl::ext::oneapi::experimental::is_property_key_of` for the property
-//     key class.
-//  5. Specialize `sycl::ext::oneapi::experimental::detail::PropertyToKind` for
-//     the new property key class. The specialization should have a `Kind`
-//     member with the value equal to the enumerator added in 1.
-//  6. Specialize
-//     `sycl::ext::oneapi::experimental::detail::IsCompileTimeProperty` for the
-//     new property key class. This specialization should derive from
-//     `std::true_type`.
-//  7. If the property needs an LLVM IR attribute, specialize
+//  4. Specialize `sycl::ext::oneapi::experimental::is_property_key_of` for the
+//     property key class.
+//  5. If the property needs an LLVM IR attribute, specialize
 //     `sycl::ext::oneapi::experimental::detail::PropertyMetaInfo` for the new
 //     `value_t` of the property key class. The specialization must have a
 //     `static constexpr const char *name` member with a value equal to the

sycl/include/sycl/ext/oneapi/properties/property.hpp

@@ -732,12 +732,28 @@ class __SYCL_TYPE(multi_ptr) multi_ptr<void, Space, DecorateAddress> {
   decorated_type *m_Pointer;
 };
 
+namespace detail {
+// See access.hpp's DecoratedType<..., access::address_space::constant_space>.
+//
+// This is only applicable to `access::decorated::legacy` mode because constant
+// AS is deprecated itself and is only accessible in legacy modes.
+template <auto Space>
+#ifdef __SYCL_DEVICE_ONLY__
+inline constexpr auto decoration_space =
+    deduce_AS<typename DecoratedType<void, Space>::type>::value;
+#else
+inline constexpr auto decoration_space = Space;
+#endif
+} // namespace detail
+
 // Legacy specialization of multi_ptr.
 // TODO: Add deprecation warning here when possible.
 template <typename ElementType, access::address_space Space>
 class __SYCL2020_DEPRECATED(
     "decorated::legacy multi_ptr specialization is deprecated since SYCL 2020.")
     multi_ptr<ElementType, Space, access::decorated::legacy> {
+  static constexpr auto DecorationSpace = detail::decoration_space<Space>;
+
 public:
   using value_type = ElementType;
   using element_type =
@@ -777,7 +793,8 @@ class __SYCL2020_DEPRECATED(
 
   multi_ptr(ElementType *pointer)
       : m_Pointer(detail::dynamic_address_cast<
-                  Space, /* SupressNotImplementedAssert = */ true>(pointer)) {
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
   }
@@ -786,7 +803,8 @@ class __SYCL2020_DEPRECATED(
   template <typename = typename detail::const_if_const_AS<Space, ElementType>>
   multi_ptr(const ElementType *pointer)
       : m_Pointer(detail::dynamic_address_cast<
-                  Space, /* SupressNotImplementedAssert = */ true>(pointer)) {}
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {}
 #endif
 
   multi_ptr(std::nullptr_t) : m_Pointer(nullptr) {}
@@ -814,7 +832,7 @@ class __SYCL2020_DEPRECATED(
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
     m_Pointer = detail::dynamic_address_cast<
-        Space, /* SupressNotImplementedAssert = */ true>(pointer);
+        DecorationSpace, /* SupressNotImplementedAssert = */ true>(pointer);
     return *this;
   }
 
@@ -856,8 +874,8 @@ class __SYCL2020_DEPRECATED(
   multi_ptr(accessor<ElementType, dimensions, Mode, target::device,
                      isPlaceholder, PropertyListT>
                 Accessor)
-      : multi_ptr(
-            detail::static_address_cast<Space>(Accessor.get_pointer().get())) {}
+      : multi_ptr(detail::static_address_cast<DecorationSpace>(
+            Accessor.get_pointer().get())) {}
 
   // Only if Space == local_space || generic_space
   template <
@@ -1088,6 +1106,8 @@ template <access::address_space Space>
 class __SYCL2020_DEPRECATED(
     "decorated::legacy multi_ptr specialization is deprecated since SYCL 2020.")
     multi_ptr<void, Space, access::decorated::legacy> {
+  static constexpr auto DecorationSpace = detail::decoration_space<Space>;
+
 public:
   using value_type = void;
   using element_type = void;
@@ -1113,17 +1133,17 @@ class __SYCL2020_DEPRECATED(
                                   !std::is_same_v<RelayPointerT, void *>>>
   multi_ptr(void *pointer)
       : m_Pointer(detail::dynamic_address_cast<
-                  Space, /* SupressNotImplementedAssert = */ true>(pointer)) {
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
   }
 #if defined(RESTRICT_WRITE_ACCESS_TO_CONSTANT_PTR)
   template <typename = typename detail::const_if_const_AS<Space, void>>
   multi_ptr(const void *pointer)
-      : m_Pointer(
-            detail::dynamic_address_cast<
-                pointer_t, /* SupressNotImplementedAssert = */ true>(pointer)) {
-  }
+      : m_Pointer(detail::dynamic_address_cast<
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {}
 #endif
 #endif
   multi_ptr(std::nullptr_t) : m_Pointer(nullptr) {}
@@ -1154,7 +1174,7 @@ class __SYCL2020_DEPRECATED(
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
     m_Pointer = detail::dynamic_address_cast<
-        Space, /* SupressNotImplementedAssert = */ true>(pointer);
+        DecorationSpace, /* SupressNotImplementedAssert = */ true>(pointer);
     return *this;
   }
 #endif
@@ -1249,6 +1269,8 @@ template <access::address_space Space>
 class __SYCL2020_DEPRECATED(
     "decorated::legacy multi_ptr specialization is deprecated since SYCL 2020.")
     multi_ptr<const void, Space, access::decorated::legacy> {
+  static constexpr auto DecorationSpace = detail::decoration_space<Space>;
+
 public:
   using value_type = const void;
   using element_type = const void;
@@ -1275,15 +1297,17 @@ class __SYCL2020_DEPRECATED(
                                   !std::is_same_v<RelayPointerT, const void *>>>
   multi_ptr(const void *pointer)
       : m_Pointer(detail::dynamic_address_cast<
-                  Space, /* SupressNotImplementedAssert = */ true>(pointer)) {
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
   }
 #if defined(RESTRICT_WRITE_ACCESS_TO_CONSTANT_PTR)
   template <typename = typename detail::const_if_const_AS<Space, void>>
   multi_ptr(const void *pointer)
       : m_Pointer(detail::dynamic_address_cast<
-                  Space, /* SupressNotImplementedAssert = */ true>(pointer)) {}
+                  DecorationSpace, /* SupressNotImplementedAssert = */ true>(
+            pointer)) {}
 #endif
 #endif
   multi_ptr(std::nullptr_t) : m_Pointer(nullptr) {}
@@ -1314,7 +1338,7 @@ class __SYCL2020_DEPRECATED(
     // TODO An implementation should reject an argument if the deduced
     // address space is not compatible with Space.
     m_Pointer = detail::dynamic_address_cast<
-        pointer_t, /* SupressNotImplementedAssert = */ true>(pointer);
+        DecorationSpace, /* SupressNotImplementedAssert = */ true>(pointer);
     return *this;
   }
 #endif
@@ -1442,7 +1466,7 @@ address_space_cast(ElementType *pointer) {
   // space is not compatible with Space.
   // Use LegacyPointerTypes here to also allow constant_space
   return multi_ptr<ElementType, Space, DecorateAddress>(
-      detail::dynamic_address_cast<Space,
+      detail::dynamic_address_cast<detail::decoration_space<Space>,
                                    /* SupressNotImplementedAssert = */ true>(
           pointer));
 }

sycl/include/sycl/multi_ptr.hpp

@@ -675,7 +675,9 @@ bool device_impl::has(aspect Aspect) const {
     const std::vector<arch> supported_archs = {
         arch::intel_cpu_spr,     arch::intel_cpu_gnr,
         arch::intel_gpu_pvc,     arch::intel_gpu_dg2_g10,
-        arch::intel_gpu_dg2_g11, arch::intel_gpu_dg2_g12};
+        arch::intel_gpu_dg2_g11, arch::intel_gpu_dg2_g12,
+        arch::intel_gpu_bmg_g21, arch::intel_gpu_lnl_m,
+        arch::intel_gpu_arl_h};
     try {
       return std::any_of(
           supported_archs.begin(), supported_archs.end(),