[NFCI][SYCL] Simplify property_key creation (#12831)

intel · Feb 28, 2024 · c0f9f23 · c0f9f23
1 parent 652f1c6
commit c0f9f23
Show file tree

Hide file tree

Showing 20 changed files with 230 additions and 878 deletions.
diff --git a/sycl/include/sycl/detail/kernel_properties.hpp b/sycl/include/sycl/detail/kernel_properties.hpp
@@ -21,7 +21,9 @@ enum class register_alloc_mode_enum : uint32_t {
  large = 2,
 };
 
-struct register_alloc_mode_key {
+struct register_alloc_mode_key
+ : ext::oneapi::experimental::detail::compile_time_property_key<
+ ext::oneapi::experimental::detail::PropKind::RegisterAllocMode> {
  template <register_alloc_mode_enum Mode>
  using value_t = sycl::ext::oneapi::experimental::property_value<
  register_alloc_mode_key,
@@ -36,19 +38,7 @@ inline constexpr register_alloc_mode_key::value_t<Mode> register_alloc_mode
 } // namespace detail
 
 namespace ext::oneapi::experimental {
-template <>
-struct is_property_key<sycl::detail::register_alloc_mode_key> : std::true_type {
-};
-
 namespace detail {
-template <> struct PropertyToKind<sycl::detail::register_alloc_mode_key> {
- static constexpr PropKind Kind = PropKind::RegisterAllocMode;
-};
-
-template <>
-struct IsCompileTimeProperty<sycl::detail::register_alloc_mode_key>
- : std::true_type {};
-
 template <sycl::detail::register_alloc_mode_enum Mode>
 struct PropertyMetaInfo<sycl::detail::register_alloc_mode_key::value_t<Mode>> {
  static constexpr const char *name = "sycl-register-alloc-mode";

diff --git a/sycl/include/sycl/ext/intel/esimd/memory_properties.hpp b/sycl/include/sycl/ext/intel/esimd/memory_properties.hpp
@@ -59,17 +59,23 @@ template <int K> inline constexpr alignment_key::value_t<K> alignment;
 /// L2 cache hint property must be used for the old/experimental LSC L3 cache
 /// hints.
 /// L3 cache property is reserved for future devices.
-struct cache_hint_L1_key {
+struct cache_hint_L1_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::ESIMDL1CacheHint> {
  template <cache_hint Hint>
  using value_t = ext::oneapi::experimental::property_value<
  cache_hint_L1_key, std::integral_constant<cache_hint, Hint>>;
 };
-struct cache_hint_L2_key {
+struct cache_hint_L2_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::ESIMDL2CacheHint> {
  template <cache_hint Hint>
  using value_t = ext::oneapi::experimental::property_value<
  cache_hint_L2_key, std::integral_constant<cache_hint, Hint>>;
 };
-struct cache_hint_L3_key {
+struct cache_hint_L3_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::ESIMDL3CacheHint> {
  template <cache_hint Hint>
  using value_t = ext::oneapi::experimental::property_value<
  cache_hint_L3_key, std::integral_constant<cache_hint, Hint>>;
@@ -151,45 +157,17 @@ struct property_value<__ESIMD_NS::cache_hint_L3_key,
  static constexpr __ESIMD_NS::cache_hint hint = Hint;
 };
 
-template <>
-struct is_property_key<sycl::ext::intel::esimd::cache_hint_L1_key>
- : std::true_type {};
-template <>
-struct is_property_key<sycl::ext::intel::esimd::cache_hint_L2_key>
- : std::true_type {};
-template <>
-struct is_property_key<sycl::ext::intel::esimd::cache_hint_L3_key>
- : std::true_type {};
-
 // Declare that esimd::properties is a property_list.
 template <typename... PropertyValueTs>
 struct is_property_list<__ESIMD_NS::properties<std::tuple<PropertyValueTs...>>>
  : is_property_list<properties<std::tuple<PropertyValueTs...>>> {};
 
 namespace detail {
-template <> struct PropertyToKind<sycl::ext::intel::esimd::cache_hint_L1_key> {
- static constexpr PropKind Kind = PropKind::ESIMDL1CacheHint;
-};
-template <> struct PropertyToKind<sycl::ext::intel::esimd::cache_hint_L2_key> {
- static constexpr PropKind Kind = PropKind::ESIMDL2CacheHint;
-};
-template <> struct PropertyToKind<sycl::ext::intel::esimd::cache_hint_L3_key> {
- static constexpr PropKind Kind = PropKind::ESIMDL3CacheHint;
-};
-
-template <>
-struct IsCompileTimeProperty<__ESIMD_NS::cache_hint_L1_key> : std::true_type {};
-template <>
-struct IsCompileTimeProperty<__ESIMD_NS::cache_hint_L2_key> : std::true_type {};
-template <>
-struct IsCompileTimeProperty<__ESIMD_NS::cache_hint_L3_key> : std::true_type {};
-
 // We do not override the class ConflictingProperties for cache_hint properties
 // because that mechanism would only allow to verify few obvious restrictions
 // without the knowledge of the context in which the cache_hint properties are
 // used (load, store, prefetch, atomic). Thus the function
 // __ESIMD_DNS::check_cache_hint() is used to verify correctness of properties.
-
 } // namespace detail
 } // namespace ext::oneapi::experimental
 } // namespace _V1

diff --git a/sycl/include/sycl/ext/intel/experimental/cache_control_properties.hpp b/sycl/include/sycl/ext/intel/experimental/cache_control_properties.hpp
@@ -72,17 +72,23 @@ template <typename PropertyT, typename... Ts>
 using property_value =
  sycl::ext::oneapi::experimental::property_value<PropertyT, Ts...>;
 
-struct read_hint_key {
+struct read_hint_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::CacheControlReadHint> {
  template <typename... Cs>
  using value_t = property_value<read_hint_key, Cs...>;
 };
 
-struct read_assertion_key {
+struct read_assertion_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::CacheControlReadAssertion> {
  template <typename... Cs>
  using value_t = property_value<read_assertion_key, Cs...>;
 };
 
-struct write_hint_key {
+struct write_hint_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::CacheControlWrite> {
  template <typename... Cs>
  using value_t = property_value<write_hint_key, Cs...>;
 };
@@ -104,21 +110,14 @@ namespace experimental {
 
 template <typename T, typename PropertyListT> class annotated_ptr;
 
-template <>
-struct is_property_key<intel::experimental::read_hint_key> : std::true_type {};
 template <typename T, typename PropertyListT>
 struct is_property_key_of<intel::experimental::read_hint_key,
  annotated_ptr<T, PropertyListT>> : std::true_type {};
 
-template <>
-struct is_property_key<intel::experimental::read_assertion_key>
- : std::true_type {};
 template <typename T, typename PropertyListT>
 struct is_property_key_of<intel::experimental::read_assertion_key,
  annotated_ptr<T, PropertyListT>> : std::true_type {};
 
-template <>
-struct is_property_key<intel::experimental::write_hint_key> : std::true_type {};
 template <typename T, typename PropertyListT>
 struct is_property_key_of<intel::experimental::write_hint_key,
  annotated_ptr<T, PropertyListT>> : std::true_type {};
@@ -178,12 +177,6 @@ template <cache_mode M> static constexpr int checkWriteHint() {
  return 0;
 }
 
-template <> struct PropertyToKind<intel::experimental::read_hint_key> {
- static constexpr PropKind Kind = PropKind::CacheControlReadHint;
-};
-template <>
-struct IsCompileTimeProperty<intel::experimental::read_hint_key>
- : std::true_type {};
 template <typename... Cs>
 struct PropertyMetaInfo<intel::experimental::read_hint_key::value_t<Cs...>> {
  static constexpr const char *name = "sycl-cache-read-hint";
@@ -196,12 +189,6 @@ struct PropertyMetaInfo<intel::experimental::read_hint_key::value_t<Cs...>> {
  ((Cs::encoding) | ...));
 };
 
-template <> struct PropertyToKind<intel::experimental::read_assertion_key> {
- static constexpr PropKind Kind = PropKind::CacheControlReadAssertion;
-};
-template <>
-struct IsCompileTimeProperty<intel::experimental::read_assertion_key>
- : std::true_type {};
 template <typename... Cs>
 struct PropertyMetaInfo<
  intel::experimental::read_assertion_key::value_t<Cs...>> {
@@ -215,12 +202,6 @@ struct PropertyMetaInfo<
  ((Cs::encoding) | ...));
 };
 
-template <> struct PropertyToKind<intel::experimental::write_hint_key> {
- static constexpr PropKind Kind = PropKind::CacheControlWrite;
-};
-template <>
-struct IsCompileTimeProperty<intel::experimental::write_hint_key>
- : std::true_type {};
 template <typename... Cs>
 struct PropertyMetaInfo<intel::experimental::write_hint_key::value_t<Cs...>> {
  static constexpr const char *name = "sycl-cache-write-hint";

diff --git a/sycl/include/sycl/ext/intel/experimental/fp_control_kernel_properties.hpp b/sycl/include/sycl/ext/intel/experimental/fp_control_kernel_properties.hpp
@@ -85,7 +85,9 @@ constexpr fp_mode setDefaultValuesIfNeeded(fp_mode mode) {
 }
 } // namespace detail
 
-struct fp_control_key {
+struct fp_control_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::FloatingPointControls> {
  template <fp_mode option>
  using value_t = ext::oneapi::experimental::property_value<
  fp_control_key, std::integral_constant<fp_mode, option>>;
@@ -97,24 +99,13 @@ inline constexpr fp_control_key::value_t<option> fp_control;
 } // namespace ext::intel::experimental
 
 namespace ext::oneapi::experimental {
-template <>
-struct is_property_key<intel::experimental::fp_control_key> : std::true_type {};
-
 template <typename T, typename PropertyListT>
 struct is_property_key_of<
  intel::experimental::fp_control_key,
  intel::experimental::kernel_attribute<T, PropertyListT>> : std::true_type {
 };
 
 namespace detail {
-template <> struct PropertyToKind<intel::experimental::fp_control_key> {
- static constexpr PropKind Kind = FloatingPointControls;
-};
-
-template <>
-struct IsCompileTimeProperty<intel::experimental::fp_control_key>
- : std::true_type {};
-
 template <intel::experimental::fp_mode FPMode>
 struct PropertyMetaInfo<intel::experimental::fp_control_key::value_t<FPMode>> {
  static_assert(intel::experimental::detail::checkMutuallyExclusive(FPMode),

diff --git a/sycl/include/sycl/ext/intel/experimental/fpga_annotated_properties.hpp b/sycl/include/sycl/ext/intel/experimental/fpga_annotated_properties.hpp
@@ -31,54 +31,68 @@ using property_value =
 //===----------------------------------------------------------------------===//
 // FPGA properties of annotated_arg/annotated_ptr
 //===----------------------------------------------------------------------===//
-struct register_map_key {
+struct register_map_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::RegisterMap> {
  using value_t = property_value<register_map_key>;
 };
 
-struct conduit_key {
+struct conduit_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::Conduit> {
  using value_t = property_value<conduit_key>;
 };
 
-struct stable_key {
+struct stable_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::Stable> {
  using value_t = property_value<stable_key>;
 };
 
-struct buffer_location_key {
+struct buffer_location_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::BufferLocation> {
  template <int K>
  using value_t =
  property_value<buffer_location_key, std::integral_constant<int, K>>;
 };
 
-struct awidth_key {
+struct awidth_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::AddrWidth> {
  template <int K>
  using value_t = property_value<awidth_key, std::integral_constant<int, K>>;
 };
 
-struct dwidth_key {
+struct dwidth_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::DataWidth> {
  template <int K>
  using value_t = property_value<dwidth_key, std::integral_constant<int, K>>;
 };
 
-struct latency_key {
+struct latency_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::Latency> {
  template <int K>
  using value_t = property_value<latency_key, std::integral_constant<int, K>>;
 };
 
 enum class read_write_mode_enum : std::uint16_t { read, write, read_write };
 
-struct read_write_mode_key {
+struct read_write_mode_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::RWMode> {
  template <read_write_mode_enum Mode>
  using value_t =
  property_value<read_write_mode_key,
  std::integral_constant<read_write_mode_enum, Mode>>;
 };
 
-struct maxburst_key {
+struct maxburst_key : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::MaxBurst> {
  template <int K>
  using value_t = property_value<maxburst_key, std::integral_constant<int, K>>;
 };
 
-struct wait_request_key {
+struct wait_request_key
+ : oneapi::experimental::detail::compile_time_property_key<
+ oneapi::experimental::detail::PropKind::WaitRequest> {
  template <int K>
  using value_t =
  property_value<wait_request_key, std::integral_constant<int, K>>;
@@ -131,18 +145,6 @@ using maxburst_key = intel::experimental::maxburst_key;
 using wait_request_key = intel::experimental::wait_request_key;
 using read_write_mode_enum = intel::experimental::read_write_mode_enum;
 
-template <> struct is_property_key<register_map_key> : std::true_type {};
-template <> struct is_property_key<conduit_key> : std::true_type {};
-template <> struct is_property_key<stable_key> : std::true_type {};
-
-template <> struct is_property_key<buffer_location_key> : std::true_type {};
-template <> struct is_property_key<awidth_key> : std::true_type {};
-template <> struct is_property_key<dwidth_key> : std::true_type {};
-template <> struct is_property_key<latency_key> : std::true_type {};
-template <> struct is_property_key<read_write_mode_key> : std::true_type {};
-template <> struct is_property_key<maxburst_key> : std::true_type {};
-template <> struct is_property_key<wait_request_key> : std::true_type {};
-
 template <typename T, typename PropertyListT>
 struct is_property_key_of<register_map_key, annotated_arg<T, PropertyListT>>
  : std::true_type {};
@@ -224,51 +226,6 @@ struct is_property_key_of<wait_request_key, annotated_ptr<T, PropertyListT>>
  : std::true_type {};
 
 namespace detail {
-template <> struct PropertyToKind<register_map_key> {
- static constexpr PropKind Kind = PropKind::RegisterMap;
-};
-template <> struct PropertyToKind<conduit_key> {
- static constexpr PropKind Kind = PropKind::Conduit;
-};
-template <> struct PropertyToKind<stable_key> {
- static constexpr PropKind Kind = PropKind::Stable;
-};
-template <> struct PropertyToKind<buffer_location_key> {
- static constexpr PropKind Kind = PropKind::BufferLocation;
-};
-template <> struct PropertyToKind<awidth_key> {
- static constexpr PropKind Kind = PropKind::AddrWidth;
-};
-template <> struct PropertyToKind<dwidth_key> {
- static constexpr PropKind Kind = PropKind::DataWidth;
-};
-template <> struct PropertyToKind<latency_key> {
- static constexpr PropKind Kind = PropKind::Latency;
-};
-template <> struct PropertyToKind<read_write_mode_key> {
- static constexpr PropKind Kind = PropKind::RWMode;
-};
-template <> struct PropertyToKind<maxburst_key> {
- static constexpr PropKind Kind = PropKind::MaxBurst;
-};
-template <> struct PropertyToKind<wait_request_key> {
- static constexpr PropKind Kind = PropKind::WaitRequest;
-};
-
-template <> struct IsCompileTimeProperty<register_map_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<conduit_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<stable_key> : std::true_type {};
-
-template <>
-struct IsCompileTimeProperty<buffer_location_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<awidth_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<dwidth_key> : std::true_type {};
-template <>
-struct IsCompileTimeProperty<read_write_mode_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<latency_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<maxburst_key> : std::true_type {};
-template <> struct IsCompileTimeProperty<wait_request_key> : std::true_type {};
-
 template <> struct PropertyMetaInfo<register_map_key::value_t> {
  static constexpr const char *name = "sycl-register-map";
  static constexpr std::nullptr_t value = nullptr;