diff --git a/sycl/include/sycl/detail/kernel_properties.hpp b/sycl/include/sycl/detail/kernel_properties.hpp index 581a2c8c5dfbf..61ecbf583d945 100644 --- 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 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 register_alloc_mode } // namespace detail namespace ext::oneapi::experimental { -template <> -struct is_property_key : std::true_type { -}; - namespace detail { -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::RegisterAllocMode; -}; - -template <> -struct IsCompileTimeProperty - : std::true_type {}; - template struct PropertyMetaInfo> { 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 index e2c9f17182fc1..4da1a64225bed 100644 --- 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 inline constexpr alignment_key::value_t 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 using value_t = ext::oneapi::experimental::property_value< cache_hint_L1_key, std::integral_constant>; }; -struct cache_hint_L2_key { +struct cache_hint_L2_key + : oneapi::experimental::detail::compile_time_property_key< + oneapi::experimental::detail::PropKind::ESIMDL2CacheHint> { template using value_t = ext::oneapi::experimental::property_value< cache_hint_L2_key, std::integral_constant>; }; -struct cache_hint_L3_key { +struct cache_hint_L3_key + : oneapi::experimental::detail::compile_time_property_key< + oneapi::experimental::detail::PropKind::ESIMDL3CacheHint> { template using value_t = ext::oneapi::experimental::property_value< cache_hint_L3_key, std::integral_constant>; @@ -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 - : std::true_type {}; -template <> -struct is_property_key - : std::true_type {}; -template <> -struct is_property_key - : std::true_type {}; - // Declare that esimd::properties is a property_list. template struct is_property_list<__ESIMD_NS::properties>> : is_property_list>> {}; namespace detail { -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::ESIMDL1CacheHint; -}; -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::ESIMDL2CacheHint; -}; -template <> struct PropertyToKind { - 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 index 4b2e4869f972c..ea8ec1020d54a 100755 --- 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 using property_value = sycl::ext::oneapi::experimental::property_value; -struct read_hint_key { +struct read_hint_key + : oneapi::experimental::detail::compile_time_property_key< + oneapi::experimental::detail::PropKind::CacheControlReadHint> { template using value_t = property_value; }; -struct read_assertion_key { +struct read_assertion_key + : oneapi::experimental::detail::compile_time_property_key< + oneapi::experimental::detail::PropKind::CacheControlReadAssertion> { template using value_t = property_value; }; -struct write_hint_key { +struct write_hint_key + : oneapi::experimental::detail::compile_time_property_key< + oneapi::experimental::detail::PropKind::CacheControlWrite> { template using value_t = property_value; }; @@ -104,21 +110,14 @@ namespace experimental { template class annotated_ptr; -template <> -struct is_property_key : std::true_type {}; template struct is_property_key_of> : std::true_type {}; -template <> -struct is_property_key - : std::true_type {}; template struct is_property_key_of> : std::true_type {}; -template <> -struct is_property_key : std::true_type {}; template struct is_property_key_of> : std::true_type {}; @@ -178,12 +177,6 @@ template static constexpr int checkWriteHint() { return 0; } -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::CacheControlReadHint; -}; -template <> -struct IsCompileTimeProperty - : std::true_type {}; template struct PropertyMetaInfo> { static constexpr const char *name = "sycl-cache-read-hint"; @@ -196,12 +189,6 @@ struct PropertyMetaInfo> { ((Cs::encoding) | ...)); }; -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::CacheControlReadAssertion; -}; -template <> -struct IsCompileTimeProperty - : std::true_type {}; template struct PropertyMetaInfo< intel::experimental::read_assertion_key::value_t> { @@ -215,12 +202,6 @@ struct PropertyMetaInfo< ((Cs::encoding) | ...)); }; -template <> struct PropertyToKind { - static constexpr PropKind Kind = PropKind::CacheControlWrite; -}; -template <> -struct IsCompileTimeProperty - : std::true_type {}; template struct PropertyMetaInfo> { 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 index 069a6fc16fb5e..544535d098e1a 100644 --- 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 using value_t = ext::oneapi::experimental::property_value< fp_control_key, std::integral_constant>; @@ -97,9 +99,6 @@ inline constexpr fp_control_key::value_t