diff --git a/sycl/doc/extensions/PropertyList/SYCL_EXT_ONEAPI_property_list.asciidoc b/sycl/doc/extensions/PropertyList/SYCL_EXT_ONEAPI_property_list.asciidoc new file mode 100644 index 0000000000000..341e023e96900 --- /dev/null +++ b/sycl/doc/extensions/PropertyList/SYCL_EXT_ONEAPI_property_list.asciidoc @@ -0,0 +1,711 @@ +:sectnums: + += `SYCL_EXT_ONEAPI_PROPERTY_LIST` + +== Introduction + +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::property_list`, which is a variant of `sycl::property_list` that supports the storage and manipulation of compile-time-constant properties in addition to runtime properties. + +== Contributors +Joe Garvey, Intel + +Roland Schulz, Intel + +Ilya Burylov, Intel + +Michael Kinsner, Intel + +John Pennycook, Intel + +Jessica Davies, Intel + +Greg Lueck, Intel + +Jason Sewall, Intel + +== Notice +Copyright (c) 2021 Intel Corporation. All rights reserved. + +== Status + +Working Draft + +This is a preview extension specification, intended to provide early access to a feature for review and community feedback. When the feature matures, this specification may be released as a formal extension. + +Because the interfaces defined by this specification are not final and are subject to change they are not intended to be used by shipping software products. + +== Version + +Last Modified Date: 2021-07-13 + +Revision: 1 + +== Dependencies + +This extension is written against the SYCL 2020 Specification (revision 3). + +== Overview + +[NOTE] +==== +In this document, we use `property_list` to indicate the proposed `sycl::ext::oneapi::property_list`. `sycl::property_list` refers to SYCL 2020, and will always include the namespace. +==== + +The purpose of this document is to clearly describe and specify `property_list` and related concepts, types, and mechanisms, and to give examples and context for their usage. + +=== Motivation + +`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 `property_list` that is able to represent compile-time-constant properties in addition to runtime properties. + +The handling of runtime properties in `property_list` 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. + +=== Examples + +[NOTE] +==== +Due to type deduction, users should usually not need to specify the template parameters of `property_list` directly. In cases it is needed `property_list_t` should be used. +==== + +A `property_list` that has the compile-time constant property `foo` can be created as follows: + +```c++ +property_list P1{foo_v}; +``` + +where `foo_v` is the global variable representing an object of the property value class of the compile-time constant `foo`. + +The type of a `property_list` is determined by the set of runtime properties and compile- +time-constant property values that it contains. In the following examples, `foo` +and `bar` are assumed to be compile-time-constant properties, while `baz` is a +runtime property. + +```c++ +property_list P1{foo_v}; +property_list P2{foo_v, baz{argv[0]}}; +property_list P3{foo_v, bar_v}; +property_list P4{bar_v, foo_v}; +static_assert(!std::is_same_v); // P1 and P2 contain different properties +static_assert(!std::is_same_v); // P1 and P3 contain different properties +static_assert(std::is_same_v); // The order properties are specified doesn't change the type +``` + +=== Goals + +The goals of this extension are: + +. Enable the storage, manipulation, and propagation of both compile-time-constant properties and runtime properties via `property_list` +. 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. + +=== Glossary + +property:: Is a class type. Properties can be used to provide extra values to classes or functions. A property has zero or more property parameters. + +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. + +property value:: An object of the property value class. +For runtime properties and properties without parameters the value type is the same as the property type. +For compile time properties with parameters the value type is given by the `value_t` type alias of the property class. + + +== Proposal + +=== Feature test macro + +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_PROPERTY_LIST` 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. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. Base features are supported. +|=== + +=== Representation of a property + +Both runtime and compile-time-constant properties are represented as class +declarations, and by convention, these classes are declared in the root of the +`sycl::ext::oneapi` namespace. The class representation of a runtime property +has no specific requirements. A runtime property typically has a constructor which takes +the value(s) of the properties and member function(s) which return those +values: + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// This is a runtime property with one integer parameter. +struct foo { + foo(int); + int value; +}; + +} // namespace oneapi +} // namespace ext +} // namespace sycl +``` + +The class representation of a compile-time-constant property must define a type +alias named `value_t`. `value_t` is templated on those parameters, and it is an alias to +an instantiation of the `property_value` class which holds the +values of the compile-time parameters. Note that the parameters to a compile- +time-constant property can be either types or non-type values. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// This property has no parameters. +struct bar { + using value_t = property_value; +}; + +// This property has one integer non-type parameter. +struct baz { + template + using value_t = property_value >; +}; + +// This property has an arbitrary number of type parameters. +struct boo { + template + using value_t = property_value; +}; + +} // namespace oneapi +} // namespace ext +} // namespace sycl +``` + +Just as with SYCL 2020 properties, all runtime and compile-time-constant +properties must have a specialization of `sycl::is_property` that inherits from +`std::true_type`, and they must have a specialization of `sycl::is_property_of` +that inherits from `std::true_type` for each SYCL runtime class that the +property can be applied to. + +```c++ +namespace sycl { + +template<> struct is_property : std::true_type {}; +template<> struct is_property : std::true_type {}; +template<> struct is_property : std::true_type {}; +template<> struct is_property : std::true_type {}; + +// These properties can be applied to any SYCL object. +template +struct is_property_of : std::true_type {}; +template +struct is_property_of : std::true_type {}; +template +struct is_property_of : std::true_type {}; +template +struct is_property_of : std::true_type {}; + +} // namespace sycl +``` + +The `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. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +template +struct property_value { + // 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 oneapi +} // namespace ext +} // namespace sycl +``` + +The members of `property_value` are described in the table below. + +-- +[options="header"] +|==== +| Member | Description +a| +```c++ +static constexpr auto value = first::value; +``` +| The value of the parameter. Available only when there is exactly one non-type +parameter. +a| +```c++ +using value_t = first; +``` +| The parameter type. Available only when there is exactly one parameter. +|=== +-- + +[NOTE] +==== +The `property_value` class is not itself a property, so it should not +specialize the `sycl::is_property` or `sycl::is_property_of` traits. +==== + + +=== Property value variables + +For each compile-time constant property a helper variable whose name has the suffix "_v" is defined. +For example, if the property is named `sycl::ext::oneapi::foo`, the pre-defined +property value variable is named `sycl::ext::oneapi::foo_v`. The variable has the type of the `value_t` alias +of the property. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// bar_v is an object of the property value type of bar. +inline constexpr bar::value_t bar_v; + +// baz_v is an object of a property value type of baz. +template +inline constexpr baz::value_t baz_v; + +// boo_v is an object of a property value type of boo. +template +inline constexpr boo::value_t boo_v; + +} // namespace oneapi +} // namespace ext +} // namespace sycl + +``` + +The implementation also provides equality and inequality operators for +properties. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// Available only if Prop is a compile-time constant property +template +constexpr bool operator==(property_value V1, property_value V2); + +// Available only if Prop is a compile-time constant property +template +constexpr bool operator!=(property_value V1, property_value V2); + +// Available only if Prop is a runtime property +template +bool operator==(Prop P1, Prop P2); + +// Available only if Prop is a runtime property +template +bool operator!=(Prop P1, Prop P2); + +} // namespace oneapi +} // namespace ext +} // namespace sycl +``` + +-- +[options="header"] +|==== +| Function | Description +a| +```c++ +template +constexpr bool operator==(property_value V1, property_value V2); +``` +| Returns true if `V1` and `V2` have the same type. Returns false otherwise. +a| +```c++ +template +constexpr bool operator!=(property_value V1, property_value V2); +``` +| Returns false if `V1` and `V2` have the same type, true otherwise. +a| +```c++ +template +bool operator==(Prop P1, Prop P2); +``` +| Returns true if all parameters (=member variables) of `P1` and `P2` are the same. Returns false otherwise. +a| +```c++ +template +bool operator!=(Prop P1, Prop P2); +``` +| Returns false if all parameters of `P1` and `P2` are the same. Returns true otherwise. +|=== +-- + + +=== `property_list` + +This extension adds a new template class, `sycl::ext::oneapi::property_list`, which is a property list that can contain compile-time constant properties as well as runtime properties. + +`sycl::ext::oneapi::property_list` is a class template and its properties influence its type. Two `sycl::ext::oneapi::property_list` 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 `sycl::ext::oneapi::property_list`, but their property values do not. +==== + +It is possible at compile-time to determine whether a `property_list` object contains a particular (runtime or compile-time constant) property. See the `static constexpr` function `has_property` of the `property_list` class. + +It is possible at compile-time to determine the property value of a compile-time constant property contained in a `property_list` object. See the `static constexpr` function `get_property` of the `property_list` class. + +The `get_property` member function of `property_list` returns the property value, which for compile-time constant properties is represented by an object of the property value class (e.g. `foo::value_t`). For runtime properties, the `get_property` member function of `property_list` returns a _copy_ of the property object passed to the `property_list` 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 `property_list`, 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 new `property_list` class template is as follows: + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +template +class property_list { + // static_assert: all types in propertiesT 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 propertiesT, + // and all types in propertiesT which are not default constructible are in propertyValueTs. + template + property_list(propertyValueTs... props); + + template + static constexpr bool has_property(); + + // Available only when propertyT is the property class of a runtime property + template + propertyT get_property() const; + + // Available only when propertyT is the property class of a compile-time constant property + template + static constexpr auto get_property(); +}; + +} // namespace oneapi +} // namespace ext +} // namespace sycl +``` + +[NOTE] +==== +Implementations will need a deduction guide to satisfy the requirement that `property_lists` which were created from the same set of compile-time constant property values and runtime properties must have an identical type. +The requirement that `property_lists` have the same type if they contain the same set of compile-time constant property values and runtime properties also implies that implementations must define an ordering over all properties. This extension provides no mechanism for users to define their own properties. +==== + +The following table describes the constructors of the `sycl::ext::oneapi::property_list` class: + +-- +[options="header"] +|==== +| Constructor | Description +a| +```c++ +template +property_list(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 propertiesT) that is not default constructable must have an object provided in props. +|=== +-- + +The following table describes the member functions of the sycl::ext::oneapi::property_list class: + +-- +[options="header"] +|==== +| Member function | Description +a| +```c++ +template +static constexpr bool has_property(); +``` | Returns true if the property list contains the property with property class propertyT. Returns false if it does not. Available only when propertyT is the property class of a (runtime or compile-time constant) property. +a| +```c++ +template +propertyT get_property() const; +``` | Returns a copy of the property value contained in the property list. +Must produce a compile error if the property_list does not contain a propertyT property. +Available only if propertyT is a runtime property. +a| +```c++ +template +static constexpr auto get_property(); +``` | Returns a copy of the property value contained in the property list. +Must produce a compile error if the property_list does not contain a propertyT property. +Available only if propertyT is the property class of a compile-time constant property. +|=== +-- + +The following trait is added to recognize a `sycl::ext::oneapi::property_list`. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// New trait to recognize a sycl::ext::oneapi::property_list +template +struct is_property_list; + +template +inline constexpr bool is_property_list_v = is_property_list::value; + +} // namespace oneapi +} // namespace ext +} // namespace sycl +``` + +The following table describes the new `is_property_list` trait: + +-- +[width="100%",options="header"] +|==== +| Traits | Description +a| +---- +template +struct is_property_list; +---- +| An explicit specialization of `is_property_list` that inherits from `std::true_type` must be provided for each specialization of `sycl::ext::oneapi::property_list`. All other specializations of `is_property_list` must inherit from `std::false_type`. +a| +---- +template +inline constexpr bool is_property_list_v; +---- +| Variable containing value of `is_property_list::value`. +|=== +-- + + +=== Type alias for property_list + +The details of the property_list template argument(s) are unspecified. In particular the sorting order of properties is unspecified. +The type of the property list can be written with the `property_list_t` type alias. The property values can be specified in any order. + +```c++ +template +using property_list_t = property_list; +``` + +The following example shows how `property_list_t` is used to create a property list containing the compile-time constant properties `bar` and `foo`: + +```c++ +using P1 = property_list_t, foo::value_t>; +using P2 = property_list_t>; +static_assert(std::is_same::value); // Succeeds, since the order of properties does not matter +static_assert(P1.get_property().value == 1); +``` + +An empty property list can be created as follows: + +```c++ +using empty_property_list = property_list_t<>; +``` + + +=== Querying Properties in Device Code + +All values of compile-time constant properties are device copyable. + +Some runtime properties may not be device copyable. +A `property_list` object is device copyable if and only if it contains no runtime properties that are not device copyable. A device copyable `property_list` 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. + +```c++ +static_assert(sycl::is_device_copyable_v)>); +static_assert(sycl::is_device_copyable_v); + +property_list P1{foo_v<1>, bar{}}; + +// All properties in P1 are device copyable, so P1 is device copyable +static_assert(sycl::is_device_copyable_v); + +h.single_task([=] { + auto a = P1.has_property(); // OK + auto b = P1.get_property(); // OK + auto c = P1.has_property(); // OK + auto d = P1.get_property(); // OK +}); +``` + +A `property_list` 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 the `property_list`, 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, otherwise the behavior is undefined. + +In the following example, `foo` is a compile-time constant property and property `bar_vec` is a runtime property that is not device copyable. + +```c++ +// 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); +std::vector v(atoi(argv[1]), 42); +property_list P2{foo_v<1>, bar_vec{v}}; +static_assert(!sycl::is_device_copyable_v); +h.single_task([=] { + auto a = decltype(P2)::has_property(); // OK, since decltype is used + auto b = P2.has_property(); // UB, since decltype is missing + + auto c = decltype(P2)::has_property(); // OK, since decltype is used + auto d = decltype(P2)::get_property(); // UB, since bar_vec is a runtime property + + auto e = decltype(P2)::get_property(); // OK, since foo is a compile-time constant property + auto f = P2.get_property(); // UB, since decltype is missing +}); +``` + +== Extended Examples + +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: + +```c++ +property_list P5{foo_v<1>}; +property_list P6{foo_v<2>}; +property_list P7{foo_v<1>, bar_v}; +static_assert(P6.has_property()); // No need to specify the value of the property's parameter +static_assert(!std::is_same_v); // The parameter values of foo are different +auto f1 = P5.get_property(); // f1 is a copy of global variable foo_v < 1 > +auto f2 = P6.get_property(); // f2 is a copy of global variable foo_v < 2 > +static_assert(f1 != f2); // Not equal since the property values are different, i.e., 1 vs. 2 +auto f3 = P7.get_property(); +static_assert(f3 == f1); // Equal because the property values are the same, i.e., equal to 1 +``` + +The 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_v` exposes the its parameters (which are types) as `first_t`, `second_t`, `third_t`. + +```c++ +property_list P8{foo_types_v}; +using f = decltype(P8.get_property()); +using t1 = f::first_t; +using t2 = f::second_t; +using t3 = f::third_t; +static_assert(std::is_same_v); +static_assert(std::is_same_v); +``` + +[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. + +```c++ +template +std::enable_if_t> my_func1(propertyListT p); + +template +std::enable_if_t && propertyListT::template has_property()> my_func2(propertyListT p); + +template +std::enable_if_t && (propertyListT::template get_property().value == 2)> my_func3(propertyListT p); +... +my_func1(property_list{foo_v}); // Legal. my_func1 accepts any properties +my_func2(property_list{foo_v}); // Legal. my_func2 requires foo +my_func2(property_list{bar_v}); // Illegal. my_func2 requires foo +my_func2(property_list{foo_v, bar_v}); // Legal. Other properties can also be specified. +my_func3(property_list{bar_v<2>); // Legal. my_func3 requires bar with value 2 +my_func3(property_list{bar_v<1>); // Illegal. my_func3 requires bar with value 2 +``` + +In the examples above, also note the distinction between the _property class_ `foo` and the _property value variable_ `foo_v`, both associated with property `foo`. Note how the property value variable `foo_v` is used to create the property list, while property class `foo` is used to query the property list. + +== End Matter + +This section provides more detailed information for implementers. It is non-normative, and may be removed in future revisions of the extension. + +=== Interface Guidelines for `property_list` consumers + +Adding a new compile-time constant property requires implementers to introduce the following: + +* A new class representing the property (referred to as the property class) +* Specializations of `sycl::is_property` and `sycl::is_property_of` for the new property class +* A global variable that represents an object of the property value + +=== Example of a Compile-time Constant Property + +This is an example showing the definition of a compile-time constant property `foo` that takes a single integer parameter. The property class associated with the property is `sycl::ext::oneapi::foo`. + +```c++ +namespace sycl { +namespace ext { +namespace oneapi { + +// foo is the property class +struct foo { + template + using value_t = property_value>; +}; + +// foo_v is a variable of the property value class that can be used to construct a +// property list with this property +template +inline constexpr foo::value_t foo_v; + +} // namespace oneapi +} // namespace ext + +// foo is a property +template<> +struct is_property : std::true_type {}; + +// foo can be applied to any object +template +struct is_property_of : std::true_type {}; + +} // namespace sycl +``` + + +== Revision History + +[cols="5,15,15,70"] +[grid="rows"] +[options="header"] +|======================================== +|Rev|Date|Author|Changes +|1|2021-07-13|Jason Sewall|Initial internal draft +|======================================== diff --git a/sycl/doc/extensions/README.md b/sycl/doc/extensions/README.md index e752bf2383f50..4697aeb8bb224 100755 --- a/sycl/doc/extensions/README.md +++ b/sycl/doc/extensions/README.md @@ -46,6 +46,7 @@ DPC++ extensions status: | [SYCL_INTEL_free_function_queries](FreeFunctionQueries/SYCL_INTEL_free_function_queries.asciidoc) | Supported (experimental) | | | [SYCL_EXT_ONEAPI_DEVICE_GLOBAL](DeviceGlobal/SYCL_INTEL_device_global.asciidoc) | Proposal | | | [SYCL_INTEL_bf16_conversion](Bf16Conversion/SYCL_INTEL_bf16_conversion.asciidoc) | Partially supported (Level Zero: GPU) | Currently available only on Xe HP GPU. ext_intel_bf16_conversion aspect is not supported. | +| [Property List](PropertyList/SYCL_EXT_ONEAPI_property_list.asciidoc) | Proposal | | Legend: