From 284f318259167c4c4d9775370729b4a5009e85e4 Mon Sep 17 00:00:00 2001 From: Konstantinos Parasyris Date: Tue, 1 Sep 2026 09:08:35 -0700 Subject: [PATCH] [SYCL] Add sycl_khr_properties container and classification traits --- sycl/include/sycl/khr/properties.hpp | 281 ++++++++++++++++++ sycl/include/sycl/sycl.hpp | 1 + .../extensions/khr_properties/properties.cpp | 96 ++++++ .../khr_properties/properties_layout.cpp | 60 ++++ .../khr_properties/properties_negative.cpp | 31 ++ 5 files changed, 469 insertions(+) create mode 100644 sycl/include/sycl/khr/properties.hpp create mode 100644 sycl/test/extensions/khr_properties/properties.cpp create mode 100644 sycl/test/extensions/khr_properties/properties_layout.cpp create mode 100644 sycl/test/extensions/khr_properties/properties_negative.cpp diff --git a/sycl/include/sycl/khr/properties.hpp b/sycl/include/sycl/khr/properties.hpp new file mode 100644 index 0000000000000..dbd7514e1cc46 --- /dev/null +++ b/sycl/include/sycl/khr/properties.hpp @@ -0,0 +1,281 @@ +//==------- properties.hpp --- sycl_khr_properties extension --------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// +// Implementation of the sycl_khr_properties extension: the +// `sycl::khr::properties` compile-time property list container and the +// associated classification traits. +// +// Design notes: +// * The list stores its properties as private base classes. Properties whose +// values are all compile-time (empty types) are NOT stored -- they are +// default-constructed on retrieval -- so an all-compile-time list is +// zero-overhead and a mixed list only pays for its runtime members. This +// keeps the list trivially copyable and cheap to compile (no std::tuple). +// * The list does not canonicalize (sort) its element order. Per the +// extension, two lists built from the same properties in a different order +// may have different types; comparison operators are intentionally not +// provided. +// +//===----------------------------------------------------------------------===// + +#pragma once + +#ifdef __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS + +#include // for __SYCL_EBO + +#include + +#define SYCL_KHR_PROPERTIES 1 + +namespace sycl { +inline namespace _V1 { +namespace khr { + +template class __SYCL_EBO properties; + +namespace detail { + +//===----------------------------------------------------------------------===// +// Base tags +// +// Every property derives (directly or via a convenience base) from +// `property_tag`, and exposes a `__detail_key_t` alias naming its key. Every +// key derives from `property_key_tag`; a key whose property has no +// runtime-provided values additionally derives from +// `compile_time_property_key_tag`. +//===----------------------------------------------------------------------===// + +struct property_tag {}; +struct property_key_tag {}; +struct compile_time_property_key_tag : property_key_tag {}; + +// Base for any property. `Key` is the property's associated key type. +template struct property_base : property_tag { + using __detail_key_t = Key; +}; + +// True if `Prop` has at least one runtime-provided value and therefore needs to +// be stored in the list (runtime and hybrid properties). A property is +// storage-free exactly when its key is a compile-time key. +template +inline constexpr bool __detail_has_runtime_value = + !std::is_base_of_v; + +//===----------------------------------------------------------------------===// +// Convenience base classes for defining properties +// +// These mirror the patterns in the extension's "Examples for implementors". +// Implementations may define properties directly on `property_base`/the tags, +// but these reduce boilerplate for the common shapes. +//===----------------------------------------------------------------------===// + +// Base for a runtime property key (all of the property's values are supplied at +// runtime). Usage: +// struct my_key : detail::runtime_property_key {}; +// struct my_prop : detail::runtime_property { int value; ... }; +struct runtime_property_key : property_key_tag {}; +template struct runtime_property : property_base {}; + +// Base for a compile-time property key with a single non-type value. Usage: +// struct my_key : detail::constant_value_property_key {}; +// template +// inline constexpr my_key::__detail_property_t my_prop; +struct constant_value_property_key : compile_time_property_key_tag { + template + struct __detail_property_t : property_base { + static constexpr Type value = Value; + }; +}; + +// Base for a compile-time property key with a single type value. Usage: +// struct my_key : detail::constant_type_property_key {}; +// template +// inline constexpr my_key::__detail_property_t my_prop; +struct constant_type_property_key : compile_time_property_key_tag { + template + struct __detail_property_t : property_base { + using value_t = Type; + }; +}; + +// Base for a hybrid property key (some values compile-time, some runtime). The +// key is a runtime key (the property carries runtime data and is stored). +// Usage: +// struct my_key : detail::hybrid_property_key {}; +// template struct my_prop : detail::hybrid_property { +// static constexpr int x = X; int y; constexpr my_prop(int y):y{y}{} }; +struct hybrid_property_key : property_key_tag {}; +template struct hybrid_property : property_base {}; + +//===----------------------------------------------------------------------===// +// Retrieval / storage machinery +//===----------------------------------------------------------------------===// + +// Selects, from a pack of properties, the one whose key is `Key`. Assumes +// exactly one match (enforced by the "no duplicate key" mandate). +template struct property_of_key; +template +struct property_of_key + : std::conditional_t, + std::enable_if, + property_of_key> {}; +template +using property_of_key_t = typename property_of_key::type; + +// Returns (by const-ref) the argument whose key is `Key`. +template +constexpr const auto &get_arg_by_key(const P &p, const Rest &...rest) { + if constexpr (std::is_same_v) + return p; + else + return get_arg_by_key(rest...); +} + +// Storage base: inherits exactly the properties that have runtime values. +// Constructed from the full property pack; each stored base is initialized from +// the matching argument. +template struct property_storage : Stored... { + constexpr property_storage() = default; + template + constexpr property_storage(const All &...all) + : Stored(get_arg_by_key(all...))... {} +}; + +// Builds `property_storage` where `subset` is the properties in `All` +// that have runtime values. +template struct build_storage; +template struct build_storage> { + using type = property_storage; +}; +template +struct build_storage, P, Rest...> + : build_storage, + property_storage, + property_storage>, + Rest...> {}; +template +using storage_for = typename build_storage, All...>::type; + +} // namespace detail + +//===----------------------------------------------------------------------===// +// Property traits +//===----------------------------------------------------------------------===// + +template +struct is_property : std::is_base_of {}; +template +inline constexpr bool is_property_v = is_property::value; + +template +struct is_property_key : std::is_base_of {}; +template +inline constexpr bool is_property_key_v = is_property_key::value; + +template +struct is_property_key_compile_time + : std::is_base_of {}; +template +inline constexpr bool is_property_key_compile_time_v = + is_property_key_compile_time::value; + +// Customization point: a property opts in for a class by specializing this +// trait for the property's key and the supported class(es). +template +struct is_property_key_for : std::false_type {}; +template +inline constexpr bool is_property_key_for_v = + is_property_key_for::value; + +// A property is for `Class` iff its key is for `Class`. +template +struct is_property_for : std::false_type {}; +template +struct is_property_for>> + : is_property_key_for {}; +template +inline constexpr bool is_property_for_v = is_property_for::value; + +// A property list all of whose properties are for `Class`. The empty list is +// for any class. +template +struct is_property_list_for : std::false_type {}; +template +struct is_property_list_for, Class> + : std::bool_constant<(is_property_for_v && ...)> {}; +template +inline constexpr bool is_property_list_for_v = + is_property_list_for::value; + +//===----------------------------------------------------------------------===// +// The properties class +//===----------------------------------------------------------------------===// + +template +class __SYCL_EBO properties + : private detail::storage_for { + using storage_t = detail::storage_for; + + static_assert((is_property_v && ...), + "Template arguments of khr::properties must be properties."); + + // Mandate: no two properties may share the same key. + template static constexpr int key_count() { + return (0 + ... + + (std::is_same_v + ? 1 + : 0)); + } + static_assert( + ((key_count() == 1) && ...), + "Duplicate properties in property list."); + +public: + template + constexpr properties(Properties... props) : storage_t(props...) {} + + template static constexpr bool has_property() { + return ((std::is_same_v) || + ...); + } + + // Compile-time key: the property carries no runtime value, so it is not + // stored; return a default-constructed instance. + template + static constexpr auto get_property() -> std::enable_if_t< + is_property_key_compile_time_v, + detail::property_of_key_t> { + return detail::property_of_key_t{}; + } + + // Runtime (or hybrid) key: return a copy of the stored property. + template + constexpr auto get_property() const -> std::enable_if_t< + !is_property_key_compile_time_v, + detail::property_of_key_t> { + return static_cast< + const detail::property_of_key_t &>( + *this); + } +}; + +// Deduction guide. Per the extension, `EncodedProperties` need not equal +// `Properties`; today it does. +template +properties(Properties... props) -> properties; + +using empty_properties_t = decltype(properties{}); + +} // namespace khr +} // namespace _V1 +} // namespace sycl + +#endif // __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS diff --git a/sycl/include/sycl/sycl.hpp b/sycl/include/sycl/sycl.hpp index 88b9065f2020b..a80341de815a9 100644 --- a/sycl/include/sycl/sycl.hpp +++ b/sycl/include/sycl/sycl.hpp @@ -166,5 +166,6 @@ can be disabled by setting SYCL_DISABLE_FSYCL_SYCLHPP_WARNING macro.") #include #include #include +#include #include #include diff --git a/sycl/test/extensions/khr_properties/properties.cpp b/sycl/test/extensions/khr_properties/properties.cpp new file mode 100644 index 0000000000000..0b3a298ea4581 --- /dev/null +++ b/sycl/test/extensions/khr_properties/properties.cpp @@ -0,0 +1,96 @@ +// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// expected-no-diagnostics +// +// Tests the sycl_khr_properties core: classification traits, the properties +// container (has_property/get_property/CTAD/empty), the feature-test macro, and +// the convenience base classes for defining runtime, compile-time, and hybrid +// properties. + +#define __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS +#include + +#ifndef SYCL_KHR_PROPERTIES +#error "SYCL_KHR_PROPERTIES feature-test macro is not defined" +#endif + +namespace kd = sycl::khr::detail; +using namespace sycl::khr; + +struct MyClass {}; +struct OtherClass {}; + +// Runtime property (convenience base + separate key). +struct enable_profiling_key : kd::runtime_property_key {}; +struct enable_profiling : kd::runtime_property { + bool value; + constexpr enable_profiling(bool v = true) : value(v) {} +}; +template <> +struct sycl::khr::is_property_key_for + : std::true_type {}; + +// Compile-time property with a single non-type value (convenience base). +struct alignment_key : kd::constant_value_property_key {}; +template +inline constexpr alignment_key::__detail_property_t + alignment; +template <> +struct sycl::khr::is_property_key_for : std::true_type { +}; + +// Compile-time property with a single type value (convenience base). +struct alignment_type_key : kd::constant_type_property_key {}; +template +inline constexpr alignment_type_key::__detail_property_t + alignment_type; +template <> +struct sycl::khr::is_property_key_for + : std::true_type {}; + +// Hybrid property (compile-time X, runtime Y) via convenience base. +struct hybrid_key : kd::hybrid_property_key {}; +template struct hybrid : kd::hybrid_property { + static constexpr int x = X; + int y; + constexpr hybrid(int y) : y(y) {} +}; +template <> +struct sycl::khr::is_property_key_for : std::true_type {}; + +// is_property / is_property_key / is_property_key_compile_time. +static_assert(is_property_v && is_property_v>); +static_assert( + is_property_v>); +static_assert(!is_property_v && !is_property_v); +static_assert(is_property_key_v && + is_property_key_v && + is_property_key_v); +static_assert(!is_property_key_v); +static_assert(is_property_key_compile_time_v && + is_property_key_compile_time_v); +// A hybrid key has runtime values, so it is NOT a compile-time key. +static_assert(!is_property_key_compile_time_v && + !is_property_key_compile_time_v); + +// is_property_key_for / is_property_for / is_property_list_for. +static_assert(is_property_key_for_v); +static_assert(!is_property_key_for_v); +static_assert(is_property_for_v && + is_property_for_v, MyClass>); +static_assert(!is_property_for_v); + +// Container: CTAD, has_property, get_property (runtime, compile-time, hybrid). +void container() { + properties p{enable_profiling{true}, alignment<16>, hybrid<3>{4}}; + static_assert(p.has_property()); + static_assert(p.has_property()); + static_assert(p.has_property()); + static_assert(!p.has_property()); + static_assert(decltype(p)::get_property().value == 16); + static_assert(!empty_properties_t::has_property()); + + // is_property_list_for over a populated and the empty list. + static_assert(is_property_list_for_v); + static_assert(!is_property_list_for_v); + static_assert(is_property_list_for_v); +} diff --git a/sycl/test/extensions/khr_properties/properties_layout.cpp b/sycl/test/extensions/khr_properties/properties_layout.cpp new file mode 100644 index 0000000000000..31a0fae374410 --- /dev/null +++ b/sycl/test/extensions/khr_properties/properties_layout.cpp @@ -0,0 +1,60 @@ +// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// expected-no-diagnostics +// +// Checks the size and trivial-copyability guarantees of khr::properties. The +// extension does not mandate these, but we ensure: compile-time-only lists are +// minimal size, mixed lists pay only for their runtime members, and lists are +// trivially copyable (unlike a std::tuple-based implementation). + +#define __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS +#include + +namespace kd = sycl::khr::detail; +using namespace sycl::khr; + +struct rt1_key : kd::runtime_property_key {}; +struct rt1 : kd::runtime_property { + bool value; + constexpr rt1(bool v = true) : value(v) {} +}; +struct rt2_key : kd::runtime_property_key {}; +struct rt2 : kd::runtime_property { + bool value; + constexpr rt2(bool v = true) : value(v) {} +}; + +struct ct1_key : kd::constant_value_property_key {}; +template +inline constexpr ct1_key::__detail_property_t ct1; +struct ct2_key : kd::constant_value_property_key {}; +template +inline constexpr ct2_key::__detail_property_t ct2; + +struct hy_key : kd::hybrid_property_key {}; +template struct hy : kd::hybrid_property { + static constexpr int x = X; + int y; + constexpr hy(int y) : y(y) {} +}; + +// Individual properties. +static_assert(sizeof(rt1{true}) == sizeof(bool)); +static_assert(sizeof(ct1<16>) == 1); +static_assert(sizeof(hy<1>{2}) == sizeof(int)); + +// Lists: compile-time-only are minimal; runtime/hybrid pay only for their +// runtime members; empty-base duplication does not bloat the list. +static_assert(sizeof(properties{}) == 1); +static_assert(sizeof(properties{ct1<16>, ct2<8>}) == 1); +static_assert(sizeof(properties{rt1{true}, rt2{false}}) == 2 * sizeof(bool)); +static_assert(sizeof(properties{rt1{true}, ct1<16>, ct2<8>}) == sizeof(bool)); +static_assert(sizeof(properties{hy<1>{2}, ct1<16>}) == sizeof(int)); + +// Trivial copyability (a std::tuple-based list would fail these). +static_assert(std::is_trivially_copyable_v); +static_assert(std::is_trivially_copyable_v>); +static_assert(std::is_trivially_copyable_v); +static_assert( + std::is_trivially_copyable_v})>); +static_assert(std::is_trivially_copyable_v{2}, rt1{true}, ct1<16>})>); diff --git a/sycl/test/extensions/khr_properties/properties_negative.cpp b/sycl/test/extensions/khr_properties/properties_negative.cpp new file mode 100644 index 0000000000000..d4615070837b9 --- /dev/null +++ b/sycl/test/extensions/khr_properties/properties_negative.cpp @@ -0,0 +1,31 @@ +// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=error,note %s +// +// Checks that misusing khr::properties produces the expected diagnostics. +// khr::properties holds its properties as private base classes, so misuse also +// produces base-class errors ("base specifier must name a class" / "base class +// specified more than once"); those are ignored here since we only assert that +// the friendly static_assert diagnostics are emitted. + +#define __DPCPP_ENABLE_UNFINISHED_KHR_EXTENSIONS +#include + +namespace kd = sycl::khr::detail; +using namespace sycl::khr; + +struct rt_key : kd::property_key_tag {}; +struct rt : kd::property_base { + int value; + constexpr rt(int v = 0) : value(v) {} +}; + +void non_property() { + // expected-error-re@sycl/khr/properties.hpp:* {{static assertion failed{{.*}}Template arguments of khr::properties must be properties.}} + properties bad{5}; + (void)bad; +} + +void duplicate_key() { + // expected-error-re@sycl/khr/properties.hpp:* {{static assertion failed{{.*}}Duplicate properties in property list.}} + properties bad{rt{1}, rt{2}}; + (void)bad; +}