The sycl::property_list found in SYCL 2020 is used to store properties used in the construction of runtime classes. It does so in a fully dynamic manner, such that it is not possible to obtain any useful information about the types of properties passed nor their values at compile time.
Compile-time-constant properties are an important building block for classes and functions that have a need to propagate compile-time information for semantic and optimization purposes, while runtime properties continue to serve an important role in dynamic parameter specification.
This extension introduces sycl::ext::oneapi::experimental::properties, which is a replacement for sycl::property_list that supports the storage and manipulation of compile-time-constant properties in addition to runtime properties.
Joe Garvey, Intel
Roland Schulz, Intel
Ilya Burylov, Intel
Michael Kinsner, Intel
John Pennycook, Intel
Jessica Davies, Intel
Greg Lueck, Intel
Jason Sewall, Intel
This is an experimental extension specification, intended to provide early access to features and gather community feedback. Interfaces defined in this specification are implemented in DPC++, but they are not finalized and may change incompatibly in future versions of DPC++ without prior notice. Shipping software products should not rely on APIs defined in this specification.
The purpose of this document is to clearly describe and specify a new property list (properties) and related concepts, types, and mechanisms, and to give examples and context for their usage.
sycl::property_list enables users to pass additional parameters when constructing objects of the runtime classes, but the design of sycl::property_list is such that it is not possible to reliably determine the contents of these passed parameters at compile time.
There are numerous circumstances where parameters are only useful when their presence and contents are known at compile-time; this extension addresses this by introducing a properties class that is able to represent compile-time-constant properties in addition to runtime properties.
The handling of runtime properties in properties departs from that of sycl::property_list by having the types of any runtime properties present be known at compile-time. The distinction between runtime properties and compile-time-constant properties is that runtime properties store values that are set at runtime.
|
Note
|
Due to type deduction, users should usually not need to specify the template parameters of |
A property list that has the compile-time constant property foo can be created as follows:
properties P1{foo};where foo is a global variable representing a property value of the compile-time constant property.
The type of a properties object is determined by the set of runtime properties and compile-
time-constant property values that it contains. In the following examples, bar
and baz are compile-time-constant properties, while foo is a
runtime property.
properties P1{bar};
properties P2{bar, foo{argv[0]}};
properties P3{bar, baz};
properties P4{baz, bar};
static_assert(!std::is_same_v<decltype(P1), decltype(P2)>); // P1 and P2 contain different properties
static_assert(!std::is_same_v<decltype(P1), decltype(P3)>); // P1 and P3 contain different properties
static_assert(std::is_same_v<decltype(P3), decltype(P4)>); // The order properties are specified doesn't change the typeThe goals of this extension are:
-
Enable the storage, manipulation, and propagation of both compile-time-constant properties and runtime properties via
properties -
Handle runtime properties such that the existence and types of runtime properties in a property list are known at compile time (while keeping their values dynamic).
The intention is to provide a robust mechanism with which to pass compile-time-constant properties and runtime properties to classes and functions.
- property
-
A property is represented by a key and value. Properties can be used to provide extra values to classes or functions.
- property value
-
An object of the property value class. A property value has zero or more property parameters. For runtime properties the value type is the same as the key type. For compile time properties the value type is given by the
value_ttype alias of the key type. - property key
-
A class representing the property key. It is used to query properties.
- property parameter
-
A parameter of a property that can be set and queried. May be dynamic (runtime) or compile-time-only, depending on the property. For compile-time-constant properties, property parameters may be types or non-types. For runtime properties, property parameters must be non-types.
- runtime property
-
A property that has property parameters determined at runtime and stored as members of the property.
- compile-time-constant property
-
A property that has no parameters that are determined at runtime. This includes properties that have no parameters and properties where all parameters are determined at compile time. If such a property has parameters, their values are stored as template arguments of the property value class.
This extension provides a feature-test macro as described in the core SYCL
specification section 6.3.3 "Feature test macros". An
implementation supporting this extension must predefine the macro
SYCL_EXT_ONEAPI_PROPERTIES to one of the values defined in the table below.
Applications can test for the existence of this macro to determine if the
implementation supports this feature, or applications can test the macro’s
value to determine which of the extension’s APIs the implementation supports.
| Value | Description |
|---|---|
1 |
Initial extension version. Base features are supported. |
Properties have a value and key type,
and by convention, these classes are declared in the root of the
sycl::ext::oneapi::experimental namespace. For a runtime property the key and value types are the same and the name of the property value
class has no suffix. A runtime property value typically has a constructor
which takes the value(s) of the properties and member function(s) which return those values.
namespace sycl::ext::oneapi::experimental {
// This is a runtime property value with one integer parameter.
// The name of the property value class is the the name of the property without any suffix.
struct foo {
foo(int);
int value;
};
// A runtime property key is an alias to the value type.
using foo_key = foo;
} // namespace experimental::oneapi::ext::syclFor compile-time constant parameters the value type is a template specialization of property_value.
The property key class contains a value_t alias which is templated on the property parameters. The property_value class holds the
values of the compile-time parameters as template arguments. The parameters to a compile-
time-constant property can be either types or non-type values.
The implementation provides a variable with the property value type. The variable has the name of the property without a suffix.
namespace sycl::ext::oneapi::experimental {
template<typename...> struct property_value;
// This property has no parameters.
struct bar_key {
using value_t = property_value<bar_key>;
};
// bar is an object of the property value type of bar.
inline constexpr bar_key::value_t bar;
// This property has one integer non-type parameter.
struct baz_key {
template<int K>
using value_t = property_value<baz_key, std::integral_constant<int, K> >;
// Note: integral_constant is used for the example. An implementation can use a different mapping.
};
// baz is an object of a property value type of baz.
template<int K>
inline constexpr baz_key::value_t<K> baz;
// This property has an arbitrary number of type parameters.
struct boo_key {
template<typename...Ts>
using value_t = property_value<boo_key, Ts...>;
};
// boo is an object of a property value type of boo.
template<typename... Ts>
inline constexpr boo_key::value_t<Ts...> boo;
} // namespace experimental::oneapi::ext::syclAll runtime and compile-time-constant properties must have a specialization of
is_property_key and is_property_value that inherits from
std::true_type, and they must have a specialization of is_property_key_of
and is_property_value_of
that inherits from std::true_type for each SYCL runtime class that the
property can be applied to. All have a base case which inherits from std::false_type.
namespace sycl::ext::oneapi::experimental {
//Base case
template<typename> struct is_property_key : std::false_type {};
template<typename, typename> struct is_property_key_of : std::false_type {};
template<> struct is_property_key<foo_key> : std::true_type {};
template<> struct is_property_key<bar_key> : std::true_type {};
// These properties can be applied to any SYCL object.
template<typename SyclObjectT>
struct is_property_key_of<foo_key, SyclObjectT> : std::true_type {};
template<typename SyclObjectT>
struct is_property_key_of<bar_key, SyclObjectT> : std::true_type {};
// is_property_value and is_property_value_of based on is_property_key(_of)
template<typename V, typename=void> struct is_property_value;
template<typename V, typename O, typename=void> struct is_property_value_of;
// Specialization for runtime properties
template<typename V> struct is_property_value<V, std::enable_if_t<(sizeof(V)>0)>> : is_property_key<V> {};
template<typename V, typename O> struct is_property_value_of<V, O, std::enable_if_t<(sizeof(V)>0)>> : is_property_key_of<V,O> {};
// Specialization for compile-time-constant properties
template<typename V> struct is_property_value<V, std::void_t<typename V::key_t>> :
is_property_key<typename V::key_t> {};
template<typename V, typename O> struct is_property_value_of<V, O, std::void_t<typename V::key_t>> :
is_property_key_of<typename V::key_t, O> {};
} // namespace experimental::oneapi::ext::syclThe property_value class has implementation-defined template parameters. In
the common case when the property has a single parameter, it provides a member
variable named value and a type alias named value_t to retrieve the value
and type of the parameter. When a property has more than one parameter, the
property_value class provides more semantically meaningful ways to retrieve
the values and types of the parameters.
namespace sycl::ext::oneapi::experimental {
template<typename Property, typename First, typename...Others>
struct property_value {
// Alias of the property key type
using key_t = Property;
// Each property with multi-parameter property_value must define template
// specializations for accessing the parameters.
// Available only when the property value has a single parameter and `value_t` of the property takes a non-type parameter
static constexpr auto value = First::value;
// Available only when the property value has a single parameter
using value_t = First;
};
} // namespace experimental::oneapi::ext::syclThe members of property_value are described in the table below.
| Member | Description |
|---|---|
static constexpr auto value = First::value; |
The value of the parameter. Available only when there is exactly one non-type parameter. |
using value_t = First; |
Available only when there is exactly one parameter. When the parameter’s value is a type, |
using key_t = property; |
The property key type. |
The implementation provides equality and inequality operators for properties.
namespace sycl::ext::oneapi::experimental {
// Available only if Prop is a compile-time constant property
template <typename Prop, typename...A, typename...B>
constexpr bool operator==(property_value<Prop, A...> V1, property_value<Prop, B...> V2);
// Available only if Prop is a compile-time constant property
template <typename Prop, typename...A, typename...B>
constexpr bool operator!=(property_value<Prop, A...> V1, property_value<Prop, B...> V2);
// Available only if Prop is a runtime property
template <typename Prop>
bool operator==(Prop P1, Prop P2);
// Available only if Prop is a runtime property
template <typename Prop>
bool operator!=(Prop P1, Prop P2);
} // namespace experimental::oneapi::ext::sycl| Function | Description |
|---|---|
template <typename Prop, typename...A, typename...B>
constexpr bool operator==(property_value<Prop, A...> V1, property_value<Prop, B...> V2); |
Returns true if |
template <typename Prop, typename...A, typename...B>
constexpr bool operator!=(property_value<Prop, A...> V1, property_value<Prop, B...> V2); |
Returns false if |
template <typename Prop>
bool operator==(Prop P1, Prop P2); |
Returns true if all parameters (=member variables) of |
template <typename Prop>
bool operator!=(Prop P1, Prop P2); |
Returns false if all parameters of |
This extension adds a new template class, sycl::ext::oneapi::experimental::properties, which is a property list that can contain compile-time constant properties as well as runtime properties.
properties is a class template, and the properties stored by it influence its type. Two properties objects have the same type if and only if they were constructed with the same set of compile-time constant property values and the same set of runtime properties.
|
Note
|
The runtime properties contained in the property list affect the type of |
It is possible at compile-time to determine whether a properties object contains a particular (runtime or compile-time constant) property. See the static constexpr function has_property of the properties class.
It is possible at compile-time to determine the property value of a compile-time constant property contained in a properties object. See the static constexpr function get_property of the properties class.
The get_property member function of properties returns a property value object.
For runtime properties, the get_property member function of properties returns a copy of the property object passed to the properties constructor.
In the same way that two different runtime properties of the same type cannot be applied to the same object, two compile-time constant property values of the same property class T cannot belong to the same properties, whether the property value is the same or different.
|
Note
|
That last sentence is not explicitly stated in the core SYCL spec, but it is assumed by the properties interface. |
The empty_properties_t type alias is a shorthand for the type of an empty property list.
The new properties class template is as follows:
namespace sycl::ext::oneapi::experimental {
template<typename PropertyValuesT>
class properties {
// static_assert: all types in PropertyValuesT need to be properties and need to be unique and sorted.
public:
// props can contain objects of compile-time constant and runtime property values in any order.
// Available only if all types in PropertyValueTs are property values.
// Only valid if all types in PropertyValueTs are in PropertyValuesT,
// and all types in PropertyValuesT which are not default constructible are in PropertyValueTs.
template<typename... PropertyValueTs>
constexpr properties(PropertyValueTs... props);
template<typename PropertyKeyT>
static constexpr bool has_property();
// Available only when PropertyT is the property class of a runtime property
template<typename PropertyT>
constexpr PropertyT get_property() const;
// Available only when PropertyKeyT is the property class of a compile-time constant property
template<typename PropertyKeyT>
static constexpr auto get_property();
};
using empty_properties_t = decltype(properties{});
} // namespace experimental::oneapi::ext::sycl|
Note
|
Implementations will need a deduction guide to satisfy the requirement that |
The following table describes the constructors of the properties class:
| Constructor | Description |
|---|---|
template<typename ... PropertyValueTs>
constexpr properties(PropertyValueTs... props); |
Available only when each argument in props is an object of a property value. Construct a property list with zero or more property values. This constructor can accept both runtime and compile-time constant property values. Each property in the property list (as determined by PropertyValuesT) that is not default constructable must have an object provided in props. |
The following table describes the member functions of the properties class:
| Member function | Description |
|---|---|
template<typename PropertyKeyT>
static constexpr bool has_property(); |
Returns true if the property list contains the property with property key class |
template<typename PropertyT>
PropertyT get_property() const; |
Returns a copy of the property value contained in the property list.
Available only if |
template<typename PropertyKeyT>
static constexpr auto get_property(); |
Returns a copy of the property value contained in the property list.
Available only if |
The following trait is added to recognize a properties.
namespace sycl::ext::oneapi::experimental {
// New trait to recognize a properties
template<typename propertyListT>
struct is_property_list;
template<typename propertyListT>
inline constexpr bool is_property_list_v = is_property_list<properties>::value;
} // namespace experimental::oneapi::ext::syclThe following table describes the new is_property_list trait:
| Traits | Description |
|---|---|
template<typename propertyListT> struct is_property_list; |
An explicit specialization of |
template<typename propertyListT> inline constexpr bool is_property_list_v; |
Variable containing value of |
The details of the properties template argument(s) are unspecified. In particular the sorting order of properties is unspecified.
The type of the property list can be written with decltype.
The following example shows how decltype is used to create a property list type containing the compile-time constant properties bar and baz:
using P1 = decltype(properties(baz<1>, bar));
using P2 = decltype(properties(bar, baz<1>));
static_assert(std::is_same<P1, P2>::value); // Succeeds, since the order of properties does not matter
static_assert(P1::get_property<baz_key>().value == 1);All values of compile-time constant properties are device copyable.
Some runtime properties may not be device copyable.
A properties object is device copyable if and only if it contains no runtime properties that are not device copyable. A device copyable properties object can be passed as a kernel parameter (as defined in the SYCL specification section 4.12.4).
In the following examples, foo is a compile-time constant property and is therefore device copyable. The property bar is a runtime property that is also device copyable.
static_assert(sycl::is_device_copyable_v<decltype(foo<1>)>);
static_assert(sycl::is_device_copyable_v<bar>);
properties P1{foo<1>, bar{}};
// All properties in P1 are device copyable, so P1 is device copyable
static_assert(sycl::is_device_copyable_v<decltype(P1)>);
h.single_task([=] {
auto a = P1.has_property<foo_key>(); // OK
auto b = P1.get_property<foo_key>(); // OK
auto c = P1.has_property<bar_key>(); // OK
auto d = P1.get_property<bar_key>(); // OK
});A properties object that contains a runtime property that is not device copyable can not be passed as a kernel parameter. However, a kernel can still call the static constexpr member functions by using decltype, as shown in the example below.
This allows a kernel to query for the existence of any property in a properties object, and it allows a kernel to query the value of a compile-time constant property, but it does not allow the kernel to query the value of a (device copyable or not device copyable) runtime property.
The following decltype syntax must be used to avoid capturing the not device copyable properties object.
In the following example, foo is a compile-time constant property and property bar_vec is a runtime property that is not device copyable.
// P2 contains the runtime property bar_vec, which is not device copyable
// P2 can not be a kernel parameter. P2 is not device copyable.
// decltype must be used even for compile-time constant properties
static_assert(!sycl::is_device_copyable_v<bar_vec>);
std::vector<int> v(atoi(argv[1]), 42);
properties P2{foo<1>, bar_vec{v}};
static_assert(!sycl::is_device_copyable_v<decltype(P2)>);
h.single_task([=] {
auto a = decltype(P2)::has_property<foo_key>(); // OK, since decltype is used
auto b = P2.has_property<foo_key>(); // incorrect, since decltype is missing
auto c = decltype(P2)::has_property<bar_vec_key>(); // OK, since decltype is used
auto d = decltype(P2)::get_property<bar_vec_key>(); // incorrect, since bar_vec is a runtime property
auto e = decltype(P2)::get_property<foo_key>(); // OK, since foo is a compile-time constant property
auto f = P2.get_property<foo_key>(); // incorrect, since decltype is missing
});The arguments passed to the constructor of a runtime property specify the value of the property’s parameters at runtime. Similarly, compile-time constant properties may have parameters that affect their semantics. For example, the property foo takes a single integer parameter:
properties P5{foo<1>};
properties P6{foo<2>};
properties P7{foo<1>, bar};
static_assert(P6.has_property<foo_key>()); // No need to specify the value of the property's parameter
static_assert(!std::is_same_v<decltype(P5), decltype(P6)>); // The parameter values of foo are different
auto f1 = P5.get_property<foo_key>(); // f1 is a copy of global variable foo<1 >
auto f2 = P6.get_property<foo_key>(); // f2 is a copy of global variable foo<2 >
static_assert(f1 != f2); // Not equal since the property values are different, i.e., 1 vs. 2
auto f3 = P7.get_property<foo_key>();
static_assert(f3 == f1); // Equal because the property values are the same, i.e., equal to 1The parameters of a property may also be types. For example, the property foo_types takes an arbitrary number of parameters, each of which is a type. In this example, foo_types exposes its parameters (which are types) as first_t, second_t, third_t.
properties P8{foo_types<float, int, bool>};
using f = decltype(P8.get_property<foo_types_key>());
using t1 = f::first_t;
using t2 = f::second_t;
using t3 = f::third_t;
static_assert(std::is_same_v<t1, float);
static_assert(std::is_same_v<t2, int>);
static_assert(std::is_same_v<t3, bool>);|
Note
|
Properties should provide semantically meaningful functions to access the parameters' values. |
For functions that take a property list parameter, it is possible to restrict
the parameter to require a specific property. If the property takes a
compile-time constant parameter, it is further possible to restrict the
function to take a property list that has a property with a specific parameter.
The following example demonstrates this, where foo is runtime property and
bar is a compile-time constant property.
template<typename PropertiesT>
std::enable_if_t<is_property_list_v<PropertiesT>> my_func1(propertyListT p);
template<typename PropertiesT>
std::enable_if_t<is_property_list_v<PropertiesT> && propertyListT::template has_property<foo_key>()> my_func2(PropertiesT p);
template<typename PropertiesT>
std::enable_if_t<is_property_list_v<PropertiesT> && (propertyListT::template get_property<bar_key>().value == 2)> my_func3(PropertiesT p);
...
my_func1(properties{foo}); // Legal. my_func1 accepts any properties
my_func2(properties{foo}); // Legal. my_func2 requires foo
my_func2(properties{bar}); // Illegal. my_func2 requires foo
my_func2(properties{foo, bar<1>}); // Legal. Other properties can also be specified.
my_func3(properties{bar<2>); // Legal. my_func3 requires bar with value 2
my_func3(properties{bar<1>); // Illegal. my_func3 requires bar with value 2In the examples above, also note the distinction between the property key class foo_key and the property value variable foo, both associated with property foo. Note how the property value variable foo is used to create the property list, while property key class foo_key is used to query the property list.
This section provides more detailed information for implementers. It is non-normative, and may be removed in future revisions of the extension.
Adding a new compile-time constant property requires implementers to introduce the following:
-
A new class representing the property key
-
Specializations of
sycl::is_property_keyandsycl::is_property_key_offor the new property class -
A global variable that represents an object of the property value
This is an example showing the definition of a compile-time constant property foo that takes a single integer parameter. The property key class associated with the property is foo_key.
namespace sycl::ext::oneapi::experimental {
// foo is the property key class
struct foo_key {
template<int K>
using value_t = property_value<foo, std::integral_constant<int, K>>;
};
// foo is a variable of the property value class that can be used to construct a
// property list with this property
template<int K>
inline constexpr foo::value_t<K> foo;
// foo is a property
template<>
struct is_property_key<foo_key> : std::true_type {};
// foo can be applied to any object
template<typename SyclObjectT>
struct is_property_key_of<foo_key, SyclObjectT> : std::true_type {};
} // namespace experimental::oneapi::ext::sycl