Skip to content

[SYCL][RTC] Rename property to registered_names #17266

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Original file line number Diff line number Diff line change
Expand Up @@ -485,53 +485,52 @@ a!
----
namespace sycl::ext::oneapi::experimental {

struct registered_kernel_names {
registered_kernel_names(); (1)
registered_kernel_names(const std::string &name); (2)
registered_kernel_names(const std::vector<std::string> &names); (3)
void add(const std::string &name); (4)
struct registered_names {
registered_names(); (1)
registered_names(const std::string &name); (2)
registered_names(const std::vector<std::string> &names); (3)
void add(const std::string &name); (4)
};
using registered_kernel_names_key = registered_kernel_names;
using registered_names_key = registered_names;

template<>
struct is_property_key<registered_kernel_names_key> : std::true_type {};
struct is_property_key<registered_names_key> : std::true_type {};

} // namespace sycl::ext::oneapi::experimental
----
!====

This property is useful when the source language represents kernel names
differently in the source code and the generated code.
For example, {cpp} function names in the generated code are "mangled" in an
implementation-defined way.
This property is useful when the source language represents names differently in
the source code and the generated code.
For example, {cpp} function names and the names of static variables at global
scope are "mangled" in an implementation-defined way in the generated code.
The precise meaning of this property is defined by each source language, but in
general it allows the application to supply a list of kernel names as they
appear in the source code.
general it allows the application to supply a list of names as they appear in
the source code.
The application can then get the corresponding raw (i.e. mangled) names after
the code is compiled.
See the section below "Obtaining a kernel when the language is ``sycl``" for a
description of how this property is used with the `source_language::sycl`
language.

_Effects (1):_ Creates a new `registered_kernel_names` property with no
registered kernel names.
_Effects (1):_ Creates a new `registered_names` property with no registered
names.

_Effects (2):_ Creates a new `registered_kernel_names` property with a single
registered kernel name.
_Effects (2):_ Creates a new `registered_names` property with a single
registered name.

_Effects (3):_ Creates a new `registered_kernel_names` property from a vector
of kernel names.
_Effects (3):_ Creates a new `registered_names` property from a vector of names.

_Effects (4):_ Adds `name` to the property's list of registered kernel names.
_Effects (4):_ Adds `name` to the property's list of registered names.

_Preconditions (2-4):_ Each source language defines its own requirements for
the registered kernel names.
_Preconditions (2-4):_ Each source language defines its own requirements for the
registered names.
For the language `source_language::sycl`, each name must be a {cpp} expression
for a pointer to a kernel function as defined below under "Obtaining a kernel
when the language is ``sycl``".

[_Note:_ It is not an error to have duplicate names in a
`registered_kernel_names` property, but the duplicates have no effect.
[_Note:_ It is not an error to have duplicate names in a `registered_names`
property, but the duplicates have no effect.
_{endnote}_]
|====

Expand Down Expand Up @@ -643,8 +642,8 @@ _Constraints:_ This function is not available when `State` is

_Returns:_ If the kernel bundle was created from a bundle of state
`bundle_state::ext_oneapi_source` and `name` was registered via
`registered_kernel_names`, returns the compiler-generated (e.g. mangled) name
for this kernel function.
`registered_names`, returns the compiler-generated (e.g. mangled) name for this
kernel function.
If the kernel bundle was created from a bundle of state
`bundle_state::ext_oneapi_source` and `name` is the same as a
compiler-generated name for a kernel defined in that bundle, that same
Expand All @@ -661,7 +660,7 @@ _Throws:_
When the kernel is defined in the language `source_language::sycl`, the host
code may query for the kernel or obtain the `kernel` object using either the
kernel's name as it is generated by the compiler (i.e. the {cpp} mangled name)
or by using the `registered_kernel_names` property.
or by using the `registered_names` property.

==== Using the compiler-generated name

Expand Down Expand Up @@ -691,11 +690,10 @@ sycl::kernel_bundle<sycl::bundle_state::executable> kb = /*...*/;
sycl::kernel k = kb.ext_oneapi_get_kernel("foo");
----

==== Using the `registered_kernel_names` property
==== Using the `registered_names` property

When the kernel is not declared as `extern "C"`, the compiler generates a
mangled name, so it is more convenient to use the `registered_kernel_names`
property.
mangled name, so it is more convenient to use the `registered_names` property.
Each string in the property must be the {cpp} expression for a pointer to a
kernel function.
These expression strings are conceptually compiled at the bottom of source
Expand Down Expand Up @@ -724,7 +722,7 @@ The host code can compile this and get the kernel's `kernel` object like so:
sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source> kb_src = /*...*/;

sycl::kernel_bundle<sycl::bundle_state::executable> kb = syclexp::build(kb_src,
syclexp::properties{syclexp::registered_kernel_names{"mykernels::bar"}});
syclexp::properties{syclexp::registered_names{"mykernels::bar"}});

sycl::kernel k = kb.ext_oneapi_get_kernel("mykernels::bar");
----
Expand All @@ -744,21 +742,21 @@ the kernel by calling `ext_oneapi_get_raw_kernel_name` like this:
sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source> kb_src = /*...*/;

sycl::kernel_bundle<sycl::bundle_state::executable> kb = syclexp::build(kb_src,
syclexp::properties{syclexp::registered_kernel_names{"mykernels::bar"}});
syclexp::properties{syclexp::registered_names{"mykernels::bar"}});

std::string mangled_name = kb.ext_oneapi_get_raw_kernel_name("mykernels::bar");
----

Again, the string passed to `ext_oneapi_get_raw_kernel_name` must have exactly
the same content as the string that was used to construct the
`registered_kernel_names` property.
the same content as the string that was used to construct the `registered_names`
property.
The application may also pass this compiler-generated (i.e. mangled) name to
`ext_oneapi_get_kernel` in order to get the `kernel` object.

==== Instantiating templated kernel functions

The `registered_kernel_names` property can also be used to instantiate a
kernel that is defined as a function template.
The `registered_names` property can also be used to instantiate a kernel that is
defined as a function template.
For example, consider source code that defines a kernel function template like
this:

Expand All @@ -774,8 +772,8 @@ std::string source = R"""(
)""";
----

The application can use the `registered_kernel_names` property to instantiate
the template for specific template arguments.
The application can use the `registered_names` property to instantiate the
template for specific template arguments.
For example, this host code instantiates the template twice and gets a `kernel`
object for each instantiation:

Expand All @@ -784,7 +782,7 @@ object for each instantiation:
sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source> kb_src = /*...*/;

sycl::kernel_bundle<sycl::bundle_state::executable> kb = syclexp::build(kb_src,
syclexp::properties{syclexp::registered_kernel_names{{"bartmpl<float>", "bartmpl<int>"}});
syclexp::properties{syclexp::registered_names{{"bartmpl<float>", "bartmpl<int>"}});

sycl::kernel k_float = kb.ext_oneapi_get_kernel("bartmpl<float>");
sycl::kernel k_int = kb.ext_oneapi_get_kernel("bartmpl<int>");
Expand Down Expand Up @@ -830,7 +828,7 @@ int main() {
syclexp::source_language::sycl,
source);

// Compile the kernel. There is no need to use the "registered_kernel_names"
// Compile the kernel. There is no need to use the "registered_names"
// property because the kernel is declared extern "C".
sycl::kernel_bundle<sycl::bundle_state::executable> kb_exe =
syclexp::build(kb_src);
Expand All @@ -852,7 +850,7 @@ int main() {

=== Disambiguating overloaded kernel functions

This example demonstrates how to use the `registered_kernel_names` property to
This example demonstrates how to use the `registered_names` property to
disambiguate a kernel function that has several overloads.

[source,c++]
Expand Down Expand Up @@ -896,11 +894,11 @@ int main() {
// use a C++ cast to disambiguate between them. Here, we are selecting the
// "int" overload.
std::string iota_name{"(void(*)(int, int*))iota"};
sycl::kernel_bundle<sycl::bundle_state::executable> kb_exe = syclexp::build(kb_src,
syclexp::properties{syclexp::registered_kernel_names{iota_name}});
sycl::kernel_bundle<sycl::bundle_state::executable> kb_exe =
syclexp::build(kb_src, syclexp::properties{syclexp::registered_names{iota_name}});

// Get the kernel by passing the same string we used to construct the
// "registered_kernel_names" property.
// "registered_names" property.
sycl::kernel iota = kb_exe.ext_oneapi_get_kernel(iota_name);

int *ptr = sycl::malloc_shared<int>(NUM, q);
Expand Down
2 changes: 1 addition & 1 deletion sycl/include/sycl/ext/oneapi/properties/property.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -210,7 +210,7 @@ enum PropKind : uint32_t {
InputDataPlacement = 65,
OutputDataPlacement = 66,
IncludeFiles = 67,
RegisteredKernelNames = 68,
RegisteredNames = 68,
ClusterLaunch = 69,
FPGACluster = 70,
Balanced = 71,
Expand Down
27 changes: 13 additions & 14 deletions sycl/include/sycl/kernel_bundle.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -1001,22 +1001,21 @@ struct is_property_key_of<save_log_key, detail::build_source_bundle_props>
: std::true_type {};

/////////////////////////
// PropertyT syclex::registered_kernel_names
// PropertyT syclex::registered_names
/////////////////////////
struct registered_kernel_names
: detail::run_time_property_key<registered_kernel_names,
detail::PropKind::RegisteredKernelNames> {
std::vector<std::string> kernel_names;
registered_kernel_names() {}
registered_kernel_names(const std::string &knArg) : kernel_names{knArg} {}
registered_kernel_names(const std::vector<std::string> &knsArg)
: kernel_names(knsArg) {}
void add(const std::string &name) { kernel_names.push_back(name); }
struct registered_names
: detail::run_time_property_key<registered_names,
detail::PropKind::RegisteredNames> {
std::vector<std::string> names;
registered_names() {}
registered_names(const std::string &name) : names{name} {}
registered_names(const std::vector<std::string> &names) : names{names} {}
void add(const std::string &name) { names.push_back(name); }
};
using registered_kernel_names_key = registered_kernel_names;
using registered_names_key = registered_names;

template <>
struct is_property_key_of<registered_kernel_names_key,
struct is_property_key_of<registered_names_key,
detail::build_source_bundle_props> : std::true_type {
};

Expand Down Expand Up @@ -1161,9 +1160,9 @@ build(kernel_bundle<bundle_state::ext_oneapi_source> &SourceKB,
if constexpr (props.template has_property<save_log>()) {
LogPtr = props.template get_property<save_log>().log;
}
if constexpr (props.template has_property<registered_kernel_names>()) {
if constexpr (props.template has_property<registered_names>()) {
RegisteredKernelNamesVec =
props.template get_property<registered_kernel_names>().kernel_names;
props.template get_property<registered_names>().names;
}
return detail::build_from_source(SourceKB, Devices, BuildOptionsVec, LogPtr,
RegisteredKernelNamesVec);
Expand Down
2 changes: 1 addition & 1 deletion sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -166,7 +166,7 @@ void test_build_and_run() {
exe_kb kbExe2 = syclex::build(
kbSrc, devs,
syclex::properties{syclex::build_options{flags}, syclex::save_log{&log},
syclex::registered_kernel_names{"ff_templated<int>"}});
syclex::registered_names{"ff_templated<int>"}});

assert(log.find("warning: 'this_nd_item<1>' is deprecated") !=
std::string::npos);
Expand Down
8 changes: 4 additions & 4 deletions sycl/test-e2e/KernelCompiler/kernel_compiler_sycl_jit.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -236,7 +236,7 @@ int test_build_and_run() {
exe_kb kbExe2 = syclex::build(
kbSrc, devs,
syclex::properties{syclex::build_options{flags}, syclex::save_log{&log},
syclex::registered_kernel_names{"ff_templated<int>"}});
syclex::registered_names{"ff_templated<int>"}});

// extern "C" was used, so the name "ff_cp" is implicitly known.
sycl::kernel k = kbExe2.ext_oneapi_get_kernel("ff_cp");
Expand Down Expand Up @@ -355,7 +355,7 @@ int test_device_code_split() {
auto build = [&](const std::string &mode) -> size_t {
exe_kb kbExe = syclex::build(
kbSrc, syclex::properties{
syclex::registered_kernel_names{names},
syclex::registered_names{names},
syclex::build_options{"-fsycl-device-code-split=" + mode}});
return std::distance(kbExe.begin(), kbExe.end());
};
Expand All @@ -372,8 +372,8 @@ int test_device_code_split() {

// Test implicit device code split
names = {"vec_add<float, 8>", "vec_add<float, 16>"};
exe_kb kbDiffWorkGroupSizes = syclex::build(
kbSrc, syclex::properties{syclex::registered_kernel_names{names}});
exe_kb kbDiffWorkGroupSizes =
syclex::build(kbSrc, syclex::properties{syclex::registered_names{names}});
assert(std::distance(kbDiffWorkGroupSizes.begin(),
kbDiffWorkGroupSizes.end()) == 2);

Expand Down
4 changes: 2 additions & 2 deletions sycl/test/abi/sycl_classes_abi_neutral_test.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -33,9 +33,9 @@
// CHECK-NEXT: 0 | class {{(std::__new_allocator|__gnu_cxx::new_allocator)}}<struct std::pair<class std::basic_string<char>, class std::basic_string<char> > > (base) (empty)
// CHECK-NEXT: 0 | {{(struct std::_Vector_base<struct std::pair<class std::basic_string<char>, class std::basic_string<char> >, class std::allocator<struct std::pair<class std::basic_string<char>, class std::basic_string<char> > > >::_Vector_impl_data \(base\)|pointer _M_start)}}

// CHECK: 0 | struct sycl::ext::oneapi::experimental::registered_kernel_names
// CHECK: 0 | struct sycl::ext::oneapi::experimental::registered_names
// CHECK-NEXT: 0 | struct sycl::ext::oneapi::experimental::detail::run_time_property_key
// CHECK: 0 | class std::vector<class std::basic_string<char> > kernel_names
// CHECK: 0 | class std::vector<class std::basic_string<char> > names
// CHECK-NEXT: 0 | struct std::_Vector_base<class std::basic_string<char>, class std::allocator<class std::basic_string<char> > > (base)
// CHECK-NEXT: 0 | struct std::_Vector_base<class std::basic_string<char>, class std::allocator<class std::basic_string<char> > >::_Vector_impl _M_impl
// CHECK-NEXT: 0 | class std::allocator<class std::basic_string<char> > (base) (empty)
Expand Down