Merge branch 'intel:sycl' into free_funtion_kernel_work_group_memory_…

…parameter

devops/dependencies-igc-dev.json

@@ -1,10 +1,10 @@
 {
   "linux": {
     "igc_dev": {
-      "github_tag": "igc-dev-6ba42ba",
-      "version": "6ba42ba",
-      "updated_at": "2024-11-13T17:40:54Z",
-      "url": "https://api.github.com/repos/intel/intel-graphics-compiler/actions/artifacts/2183383213/zip",
+      "github_tag": "igc-dev-0b4b682",
+      "version": "0b4b682",
+      "updated_at": "2024-11-17T01:09:50Z",
+      "url": "https://api.github.com/repos/intel/intel-graphics-compiler/actions/artifacts/2197388704/zip",
       "root": "{DEPS_ROOT}/opencl/runtime/linux/oclgpu"
     }
   }

sycl/include/sycl/ext/intel/experimental/fpga_annotated_properties.hpp

@@ -319,28 +319,25 @@ struct propagateToPtrAnnotation<buffer_location_key> : std::true_type {};
 //===----------------------------------------------------------------------===//
 //
 namespace detail {
-template <typename... Args> struct checkValidFPGAPropertySet {
-  using list = std::tuple<Args...>;
-  static constexpr bool has_BufferLocation =
-      ContainsProperty<buffer_location_key, list>::value;
+template <typename property_list_t> struct checkValidFPGAPropertySet {
+  template <typename... Keys>
+  static constexpr bool has_one_of =
+      ((property_list_t::template has_property<Keys>() || ...));
+
+  static constexpr bool has_BufferLocation = has_one_of<buffer_location_key>;
 
   static constexpr bool has_InterfaceConfig =
-      ContainsProperty<awidth_key, list>::value ||
-      ContainsProperty<dwidth_key, list>::value ||
-      ContainsProperty<latency_key, list>::value ||
-      ContainsProperty<read_write_mode_key, list>::value ||
-      ContainsProperty<maxburst_key, list>::value ||
-      ContainsProperty<wait_request_key, list>::value;
+      has_one_of<awidth_key, dwidth_key, latency_key, read_write_mode_key,
+                 maxburst_key, wait_request_key>;
 
   static constexpr bool value = !(!has_BufferLocation && has_InterfaceConfig);
 };
 
-template <typename... Args> struct checkHasConduitAndRegisterMap {
-  using list = std::tuple<Args...>;
+template <typename property_list_t> struct checkHasConduitAndRegisterMap {
   static constexpr bool has_Conduit =
-      ContainsProperty<conduit_key, list>::value;
+      property_list_t::template has_property<conduit_key>();
   static constexpr bool has_RegisterMap =
-      ContainsProperty<register_map_key, list>::value;
+      property_list_t::template has_property<register_map_key>();
   static constexpr bool value = !(has_Conduit && has_RegisterMap);
 };
 } // namespace detail

sycl/include/sycl/ext/oneapi/experimental/annotated_arg/annotated_arg.hpp

@@ -214,13 +214,13 @@ __SYCL_TYPE(annotated_arg) annotated_arg<T *, detail::properties_t<Props...>> {
                 "The property list contains invalid property.");
   // check the set if FPGA specificed properties are used
   static constexpr bool hasValidFPGAProperties =
-      detail::checkValidFPGAPropertySet<Props...>::value;
+      detail::checkValidFPGAPropertySet<property_list_t>::value;
   static_assert(hasValidFPGAProperties,
                 "FPGA Interface properties (i.e. awidth, dwidth, etc.) "
                 "can only be set with BufferLocation together.");
   // check if conduit and register_map properties are specified together
   static constexpr bool hasConduitAndRegisterMapProperties =
-      detail::checkHasConduitAndRegisterMap<Props...>::value;
+      detail::checkHasConduitAndRegisterMap<property_list_t>::value;
   static_assert(hasConduitAndRegisterMapProperties,
                 "The properties conduit and register_map cannot be "
                 "specified at the same time.");
@@ -447,13 +447,13 @@ __SYCL_TYPE(annotated_arg) annotated_arg<T, detail::properties_t<Props...>> {
                 "The property list contains invalid property.");
   // check the set if FPGA specificed properties are used
   static constexpr bool hasValidFPGAProperties =
-      detail::checkValidFPGAPropertySet<Props...>::value;
+      detail::checkValidFPGAPropertySet<property_list_t>::value;
   static_assert(hasValidFPGAProperties,
                 "FPGA Interface properties (i.e. awidth, dwidth, etc.) "
                 "can only be set with BufferLocation together.");
   // check if conduit and register_map properties are specified together
   static constexpr bool hasConduitAndRegisterMapProperties =
-      detail::checkHasConduitAndRegisterMap<Props...>::value;
+      detail::checkHasConduitAndRegisterMap<property_list_t>::value;
   static_assert(hasConduitAndRegisterMapProperties,
                 "The properties conduit and register_map cannot be "
                 "specified at the same time.");

sycl/include/sycl/ext/oneapi/experimental/annotated_ptr/annotated_ptr.hpp

@@ -50,10 +50,6 @@ template <class T>
 constexpr bool is_ann_ref_v =
     is_ann_ref_impl<std::remove_reference_t<T>>::value;
 
-template <typename... Ts>
-using contains_alignment =
-    detail::ContainsProperty<alignment_key, std::tuple<Ts...>>;
-
 // filter properties that are applied on annotations
 template <typename PropertyListTy>
 using annotation_filter = decltype(filter_properties<propagateToPtrAnnotation>(
@@ -392,69 +388,69 @@ __SYCL_TYPE(annotated_ptr) annotated_ptr<T, detail::properties_t<Props...>> {
   // turned off for these operators to make sure the complete error notes are
   // printed
   // clang-format off
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   reference operator[](std::ptrdiff_t idx) const noexcept {
     return reference(m_Ptr + idx);
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator[](std::ptrdiff_t idx) const noexcept -> decltype("operator[] is not available when alignment is specified!") = delete;
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   annotated_ptr operator+(size_t offset) const noexcept {
     return annotated_ptr(m_Ptr + offset);
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator+(size_t offset) const noexcept -> decltype("operator+ is not available when alignment is specified!") = delete;
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   annotated_ptr &operator++() noexcept {
     m_Ptr += 1;
     return *this;
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator++() noexcept -> decltype("operator++ is not available when alignment is specified!") = delete;
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   annotated_ptr operator++(int) noexcept {
     auto tmp = *this;
     m_Ptr += 1;
     return tmp;
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator++(int) noexcept -> decltype("operator++ is not available when alignment is specified!") = delete;
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   annotated_ptr &operator--() noexcept {
     m_Ptr -= 1;
     return *this;
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator--() noexcept -> decltype("operator-- is not available when alignment is specified!") = delete;
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<!has_alignment>>
   annotated_ptr operator--(int) noexcept {
     auto tmp = *this;
     m_Ptr -= 1;
     return tmp;
   }
 
-  template <bool has_alignment = detail::contains_alignment<Props...>::value,
+  template <bool has_alignment = property_list_t::template has_property<alignment_key>(),
             class = std::enable_if_t<has_alignment>>
   auto operator--(int) noexcept -> decltype("operator-- is not available when alignment is specified!") = delete;
 
@@ -485,13 +481,13 @@ __SYCL_TYPE(annotated_ptr) annotated_ptr<T, detail::properties_t<Props...>> {
                 "The property list contains invalid property.");
   // check the set if FPGA specificed properties are used
   static constexpr bool hasValidFPGAProperties =
-      detail::checkValidFPGAPropertySet<Props...>::value;
+      detail::checkValidFPGAPropertySet<property_list_t>::value;
   static_assert(hasValidFPGAProperties,
                 "FPGA Interface properties (i.e. awidth, dwidth, etc.) "
                 "can only be set with BufferLocation together.");
   // check if conduit and register_map properties are specified together
   static constexpr bool hasConduitAndRegisterMapProperties =
-      detail::checkHasConduitAndRegisterMap<Props...>::value;
+      detail::checkHasConduitAndRegisterMap<property_list_t>::value;
   static_assert(hasConduitAndRegisterMapProperties,
                 "The properties conduit and register_map cannot be "
                 "specified at the same time.");

sycl/include/sycl/ext/oneapi/experimental/annotated_ptr/annotated_ptr_properties.hpp

@@ -55,24 +55,26 @@ struct PropertyMetaInfo<usm_kind_key::value_t<Kind>> {
   static constexpr sycl::usm::alloc value = Kind;
 };
 
-template <typename PropertyListT> struct IsUsmKindDevice : std::false_type {};
-template <typename... Props>
-struct IsUsmKindDevice<detail::properties_t<Props...>>
-    : detail::ContainsProperty<std::remove_const_t<decltype(usm_kind_device)>,
-                               std::tuple<Props...>> {};
-
-template <typename PropertyListT> struct IsUsmKindHost : std::false_type {};
-template <typename... Props>
-struct IsUsmKindHost<detail::properties_t<Props...>>
-    : detail::ContainsProperty<std::remove_const_t<decltype(usm_kind_host)>,
-                               std::tuple<Props...>> {};
-
-template <typename PropertyListT> struct IsUsmKindShared : std::false_type {};
-template <typename... Props>
-struct IsUsmKindShared<detail::properties_t<Props...>>
-    : detail::ContainsProperty<std::remove_const_t<decltype(usm_kind_shared)>,
-                               std::tuple<Props...>> {};
+template <typename PropertyListT, sycl::usm::alloc Kind>
+inline constexpr bool is_usm_kind = []() constexpr {
+  if constexpr (PropertyListT::template has_property<usm_kind_key>())
+    return PropertyListT::template get_property<usm_kind_key>() ==
+           usm_kind<Kind>;
+  else
+    return false;
+}();
 
+template <typename PropertyListT>
+struct IsUsmKindDevice
+    : std::bool_constant<is_usm_kind<PropertyListT, sycl::usm::alloc::device>> {
+};
+template <typename PropertyListT>
+struct IsUsmKindHost
+    : std::bool_constant<is_usm_kind<PropertyListT, sycl::usm::alloc::host>> {};
+template <typename PropertyListT>
+struct IsUsmKindShared
+    : std::bool_constant<is_usm_kind<PropertyListT, sycl::usm::alloc::shared>> {
+};
 } // namespace detail
 
 } // namespace experimental

sycl/include/sycl/ext/oneapi/experimental/annotated_usm/alloc_util.hpp

@@ -32,11 +32,9 @@ using MergeUsmKind =
                                 decltype(properties{usm_kind<Kind>})>;
 
 // Check if a property list contains the a certain property
-template <typename PropKey, typename PropertyListT> struct HasProperty {};
-
-template <typename PropKey, typename... Props>
-struct HasProperty<PropKey, detail::properties_t<Props...>>
-    : detail::ContainsProperty<PropKey, std::tuple<Props...>> {};
+template <typename PropKey, typename PropertyListT>
+struct HasProperty
+    : std::bool_constant<PropertyListT::template has_property<PropKey>()> {};
 
 template <typename PropertyListT>
 using HasAlign = HasProperty<alignment_key, PropertyListT>;
@@ -91,13 +89,13 @@ struct ValidAllocPropertyList<T, detail::properties_t<Prop, Props...>>
                  IsRuntimePropertyValid<Prop>::value) &&
                     ValidAllocPropertyList<
                         T, detail::properties_t<Props...>>::value> {
+  static_assert(is_property_value_v<Prop>);
+  static constexpr bool is_compile_time = std::is_empty_v<Prop>;
   // check if a compile-time property is valid for annotated_ptr
-  static_assert(!detail::IsCompileTimePropertyValue<Prop>::value ||
-                    is_valid_property<T *, Prop>::value,
+  static_assert(!is_compile_time || is_valid_property<T *, Prop>::value,
                 "Found invalid compile-time property in the property list.");
   // check if a runtime property is valid for malloc
-  static_assert(!detail::IsRuntimeProperty<Prop>::value ||
-                    IsRuntimePropertyValid<Prop>::value,
+  static_assert(is_compile_time || IsRuntimePropertyValid<Prop>::value,
                 "Found invalid runtime property in the property list.");
 };
 
@@ -112,15 +110,15 @@ template <> struct GetCompileTimeProperties<empty_properties_t> {
 template <typename Prop>
 struct GetCompileTimeProperties<detail::properties_t<Prop>> {
   using type =
-      std::conditional_t<detail::IsCompileTimePropertyValue<Prop>::value,
-                         detail::properties_t<Prop>, empty_properties_t>;
+      std::conditional_t<std::is_empty_v<Prop>, detail::properties_t<Prop>,
+                         empty_properties_t>;
 };
 
 template <typename Prop, typename... Props>
 struct GetCompileTimeProperties<detail::properties_t<Prop, Props...>> {
   using filtered_this_property_t =
-      std::conditional_t<detail::IsCompileTimePropertyValue<Prop>::value,
-                         detail::properties_t<Prop>, empty_properties_t>;
+      std::conditional_t<std::is_empty_v<Prop>, detail::properties_t<Prop>,
+                         empty_properties_t>;
   using filtered_other_properties_t =
       typename GetCompileTimeProperties<detail::properties_t<Props...>>::type;
   using type = detail::merged_properties_t<filtered_this_property_t,

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

@@ -289,29 +289,6 @@ template <typename PropertyT> struct PropertyID {
       static_cast<int>(PropertyToKind<PropertyT>::Kind);
 };
 
-// Trait for identifying runtime properties.
-template <typename PropertyT>
-struct IsRuntimeProperty
-    : std::bool_constant<
-          !is_property_list_v<PropertyT> &&
-          std::is_base_of_v<property_key_base_tag, PropertyT> &&
-          !std::is_base_of_v<compile_time_property_key_base_tag, PropertyT>> {};
-
-// Trait for identifying compile-time properties.
-template <typename PropertyT>
-struct IsCompileTimeProperty
-    : std::bool_constant<
-          !is_property_list_v<PropertyT> &&
-          std::is_base_of_v<property_key_base_tag, PropertyT> &&
-          std::is_base_of_v<compile_time_property_key_base_tag, PropertyT>> {};
-
-// Checks if a type is either a runtime property or if it is a compile-time
-// property
-template <typename T> struct IsProperty {
-  static constexpr bool value =
-      IsRuntimeProperty<T>::value || IsCompileTimeProperty<T>::value;
-};
-
 // Trait for property compile-time meta names and values.
 template <typename PropertyT> struct PropertyMetaInfo {
   // Some properties don't have meaningful compile-time values.

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

@@ -43,20 +43,6 @@ template <int N, typename... Ts>
 using nth_type_t = typename nth_type<N, Ts...>::type;
 #endif
 
-//******************************************************************************
-// Property identification
-//******************************************************************************
-
-// Checks that all types in a tuple are valid properties.
-template <typename T> struct AllPropertyValues {};
-template <typename... Ts>
-struct AllPropertyValues<std::tuple<Ts...>> : std::true_type {};
-template <typename T, typename... Ts>
-struct AllPropertyValues<std::tuple<T, Ts...>>
-    : std::conditional_t<IsPropertyValue<T>::value,
-                         AllPropertyValues<std::tuple<Ts...>>,
-                         std::false_type> {};
-
 //******************************************************************************
 // Property value tooling
 //******************************************************************************