Skip to content

[SYCL][Docs] Add proposed event mode extension #15704

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.

Sign up for GitHub

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 8 commits into from
Oct 31, 2024
Merged

[SYCL][Docs] Add proposed event mode extension #15704

Show file tree
Hide file tree
Changes from 3 commits
Commits
Show all changes
8 commits
Select commit Hold shift + click to select a range
9110431
[SYCL][Docs] Add proposed low-power event extension
steffenlarsen Oct 15, 2024
7c45672
Change to properties on submit
steffenlarsen Oct 16, 2024
19cdc17
Update sycl/doc/extensions/proposed/sycl_ext_intel_low_power_event.as…
steffenlarsen Oct 16, 2024
276cc80
Move properties argument and fix editorial issues
steffenlarsen Oct 17, 2024
56129d1
Fix new overloads
steffenlarsen Oct 17, 2024
f96675e
Switch to event-mode property
steffenlarsen Oct 22, 2024
7cf93b4
Update sycl_ext_intel_event_mode.asciidoc
steffenlarsen Oct 23, 2024
197d433
Switch to event_mode_enum
steffenlarsen Oct 29, 2024
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
9 changes: 5 additions & 4 deletions sycl/doc/extensions/experimental/sycl_ext_oneapi_enqueue_functions.asciidoc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -230,8 +230,8 @@ a!
----
namespace sycl::ext::oneapi::experimental {

template <typename CommandGroupFunc>
void submit(sycl::queue q, CommandGroupFunc&& cgf);
template <typename CommandGroupFunc, typename Properties = empty_properties_t>
void submit(sycl::queue q, CommandGroupFunc&& cgf, Properties properties = {});

}
----
Expand All @@ -247,8 +247,9 @@ a!
----
namespace sycl::ext::oneapi::experimental {

template <typename CommandGroupFunc>
sycl::event submit_with_event(sycl::queue q, CommandGroupFunc&& cgf);
template <typename CommandGroupFunc, typename Properties = empty_properties_t>
sycl::event submit_with_event(sycl::queue q, CommandGroupFunc&& cgf,
Properties properties = {});

}
----
Expand Down
156 changes: 156 additions & 0 deletions sycl/doc/extensions/proposed/sycl_ext_intel_low_power_event.asciidoc
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,156 @@
= sycl_ext_intel_low_power_event

:source-highlighter: coderay
:coderay-linenums-mode: table

// This section needs to be after the document title.
:doctype: book
:toc2:
:toc: left
:encoding: utf-8
:lang: en
:dpcpp: pass:[DPC++]
:endnote: &#8212;{nbsp}end{nbsp}note

// Set the default source code type in this document to C++,
// for syntax highlighting purposes. This is needed because
// docbook uses c++ and html5 uses cpp.
:language: {basebackend@docbook:c++:cpp}

:common_ref_sem: https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#sec:reference-semantics

== Notice

[%hardbreaks]
Copyright (C) 2024 Intel Corporation. All rights reserved.

Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks
of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by
permission by Khronos.


== Contact

To report problems with this extension, please open a new issue at:

https://github.com/intel/llvm/issues


== Dependencies

This extension is written against the SYCL 2020 revision 9 specification. All
references below to the "core SYCL specification" or to section numbers in the
SYCL specification refer to that revision.

This extension also depends on the following other SYCL extensions:

* link:../experimental/sycl_ext_oneapi_enqueue_functions.asciidoc[
sycl_ext_oneapi_enqueue_functions]
* link:../experimental/sycl_ext_oneapi_properties.asciidoc[
sycl_ext_oneapi_properties]


== Status

This is a proposed extension specification, intended to gather community
feedback. Interfaces defined in this specification may not be implemented yet
or may be in a preliminary state. The specification itself may also change in
incompatible ways before it is finalized. *Shipping software products should
not rely on APIs defined in this specification.*


== Backend support status

This extension is currently implemented in {dpcpp} only for all device targets.

== Overview

On some backends, calling `wait()` on a `event` will synchronize using a
busy-waiting implementation. Though this comes at a low latency for the
synchronization of the event, it has the downside of consuming high amounts of
CPU time for no meaningful work. This extension introduces a new property for
SYCL commands that will produce a "low-power" event. These new low-power events
will, if possible, yield the thread that the `wait()` member function is called
on and only wake up occasionally to check if the event has finished. This
reduces the time the CPU spends checking finish condition of the wait, at the
cost of latency.


== Specification

=== Feature test macro

This extension provides a feature-test macro as described in the core SYCL
specification. An implementation supporting this extension must predefine the
macro `SYCL_EXT_ONEAPI_LOW_POWER_EVENT` 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 features the implementation
supports.

[%header,cols="1,5"]
|===
|Value
|Description

|1
|The APIs of this experimental extension are not versioned, so the
feature-test macro always has this value.
|===


=== Low-power event property

This extension adds a new property which can be used with the
`submit_with_event` free function from
link:../experimental/sycl_ext_oneapi_enqueue_functions.asciidoc[sycl_ext_oneapi_enqueue_functions].
Passing this property to either of these functions will act as a hint to the
`event` created from the corresponding commands to do low-power synchronization.
If the backend is able to handle low-power events, calling `event::wait()` or
`event::wait_and_throw()` will cause the thread to yield and only do occasional
wake-ups to check the event progress.

[_Note:_ The property currently only has an effect on `barrier` and
`partial_barrier` commands enqueued on queues that return
`backend::ext_oneapi_level_zero` from `queue::get_backend()`.
_{endnote}_]

```
namespace sycl::ext::intel::experimental {

struct low_power_event_key {
using value_t =
oneapi::experimental::property_value<low_power_event_key>;
};

inline constexpr low_power_event_key::value_t low_power_event;
Copy link
Contributor

@gmlueck gmlueck Oct 16, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think this should be a runtime property that takes a bool parameter. This will allow usage like:

bool low_power = /* check some option */;
oneapiex::properties props{intelex::low_power_event{low_power}});

Code like this is difficult with the API as you have it now because the existence of the low_power_event property affects the type of the props variable. As a result, you can not do this:

bool low_power = /* check some option */;
oneapiex::properties props;
if (low_power)
  /* no way to add 'low_power_event' property now */

If you make it a runtime property, the property's default constructor can set the property's value to true.

Alternatively, the property could take an enum instead of a bool. I think someone suggested in one of our meetings that there could be other types of low-power events in the future? I can't remember whether we decided that was realistic, though.

Copy link
Contributor Author

@steffenlarsen steffenlarsen Oct 17, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

That's a good point. Would it make sense to change the property to an event_mode property and have the value as:

enum event_mode_enum {
  default_mode,
  low_power_mode
};

If we think future modes might overlap, we can also do it as a bitmap.

Copy link
Contributor

@Pennycook Pennycook Oct 17, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Am I right in thinking that we can define it as an enum now, and convert it to a bitmask later just by defining the relevant operators?

Either way, I like this direction. I'd be tempted to change the names, though, to avoid some redundancy:

enum class event_mode {
  none, // we can't use"default" because it's a keyword
  low_power,
};

Copy link
Contributor Author

@steffenlarsen steffenlarsen Oct 17, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Am I right in thinking that we can define it as an enum now, and convert it to a bitmask later just by defining the relevant operators?

Absolutely. As long as it doesn't cross the library boundary, we can change it later if we want.

Copy link
Contributor

@gmlueck gmlueck Oct 17, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think the enum direction makes sense, but I'm not sure what names you are proposing. It seems like the name event_mode has been proposed as both the name of the property and the name of the enumeration. Can you write out an example usage to illustrate the naming that you propose?

Copy link
Contributor Author

@steffenlarsen steffenlarsen Oct 18, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I'm definitely not married to the name of the enum, but I just took inspiration from the naming of the compile-time property enums.

Adjusting the example in the text, the property would be used as

oneapiex::properties Props{intelex::event_mode{intelex::event_mode_enum::low_power}};

Note, I agree with @Pennycook in that it should be an enum class, mainly so it doesn't bleed its contents.


} // namespace sycl::ext::intel::experimental
```

=== New property usage example

As an example of how to use the new `low_power_event` property, see the
following code:

```
#include <sycl/sycl.hpp>

namespace oneapiex = sycl::ext::oneapi::experimental;
namespace intelex = sycl::ext::intel::experimental;

int main() {
sycl::queue Q;

// Submit some work to the queue.
oneapiex::submit(Q, [&](sycl::handler &CGH) {...});

// Submit a command with the low-power event property.
sycl::event E = oneapiex::submit_with_event(Q, [&](sycl::handler &CGH) {
...
}, oneapiex::properties{intelex::low_power_event});

// Waiting for the resulting event will use low-power waiting if possible.
E.wait();
}
```
10 changes: 6 additions & 4 deletions sycl/include/sycl/ext/oneapi/experimental/enqueue_functions.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -95,18 +95,20 @@ void submit_impl(queue &Q, CommandGroupFunc &&CGF,
}
} // namespace detail

template <typename CommandGroupFunc>
void submit(queue Q, CommandGroupFunc &&CGF,
template <typename CommandGroupFunc, typename PropertiesT = empty_properties_t>
void submit(queue Q, CommandGroupFunc &&CGF, PropertiesT Props = {},
const sycl::detail::code_location &CodeLoc =
sycl::detail::code_location::current()) {
std::ignore = Props;
sycl::ext::oneapi::experimental::detail::submit_impl(
Q, std::forward<CommandGroupFunc>(CGF), CodeLoc);
}

template <typename CommandGroupFunc>
event submit_with_event(queue Q, CommandGroupFunc &&CGF,
template <typename CommandGroupFunc, typename PropertiesT = empty_properties_t>
event submit_with_event(queue Q, CommandGroupFunc &&CGF, PropertiesT Props = {},
const sycl::detail::code_location &CodeLoc =
sycl::detail::code_location::current()) {
std::ignore = Props;
return Q.submit(std::forward<CommandGroupFunc>(CGF), CodeLoc);
}

Expand Down
Loading