Skip to content

Commit 1600797

Browse files
steffenlarsensteffenlarsen
committed
[SYCL] Implement sycl_ext_oneapi_event_mode extension
This commit implements the sycl_ext_oneapi_event_mode extension. Of particular focus is the low-power event mode. Signed-off-by: Larsen, Steffen <[email protected]>
1 parent 528d43a commit 1600797

25 files changed

+380
-79
lines changed

‎sycl/include/sycl/ext/oneapi/experimental/enqueue_functions.hpp‎

Lines changed: 14 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -89,20 +89,27 @@ template <typename LCRangeT, typename LCPropertiesT> struct LaunchConfigAccess {
8989
}
9090
};
9191

92-
template <typename CommandGroupFunc>
93-
void submit_impl(queue &Q, CommandGroupFunc &&CGF,
92+
template <typename CommandGroupFunc, typename PropertiesT>
93+
void submit_impl(queue &Q, PropertiesT Props, CommandGroupFunc &&CGF,
9494
const sycl::detail::code_location &CodeLoc) {
95-
Q.submit_without_event(std::forward<CommandGroupFunc>(CGF), CodeLoc);
95+
Q.submit_without_event(Props, std::forward<CommandGroupFunc>(CGF), CodeLoc);
96+
}
97+
98+
template <typename CommandGroupFunc, typename PropertiesT>
99+
event submit_with_event_impl(queue &Q, PropertiesT Props,
100+
CommandGroupFunc &&CGF,
101+
const sycl::detail::code_location &CodeLoc) {
102+
return Q.submit_with_event(Props, std::forward<CommandGroupFunc>(CGF),
103+
nullptr, CodeLoc);
96104
}
97105
} // namespace detail
98106

99107
template <typename CommandGroupFunc, typename PropertiesT>
100108
void submit(queue Q, PropertiesT Props, CommandGroupFunc &&CGF,
101109
const sycl::detail::code_location &CodeLoc =
102110
sycl::detail::code_location::current()) {
103-
std::ignore = Props;
104111
sycl::ext::oneapi::experimental::detail::submit_impl(
105-
Q, std::forward<CommandGroupFunc>(CGF), CodeLoc);
112+
Q, Props, std::forward<CommandGroupFunc>(CGF), CodeLoc);
106113
}
107114

108115
template <typename CommandGroupFunc>
@@ -116,8 +123,8 @@ template <typename CommandGroupFunc, typename PropertiesT>
116123
event submit_with_event(queue Q, PropertiesT Props, CommandGroupFunc &&CGF,
117124
const sycl::detail::code_location &CodeLoc =
118125
sycl::detail::code_location::current()) {
119-
std::ignore = Props;
120-
return Q.submit(std::forward<CommandGroupFunc>(CGF), CodeLoc);
126+
return sycl::ext::oneapi::experimental::detail::submit_with_event_impl(
127+
Q, Props, std::forward<CommandGroupFunc>(CGF), CodeLoc);
121128
}
122129

123130
template <typename CommandGroupFunc>

‎sycl/include/sycl/ext/oneapi/experimental/event_mode_property.hpp‎

Lines changed: 35 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,35 @@
1+
//==-- cluster_group_prop.hpp --- SYCL extension for event mode property ---==//
2+
//
3+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4+
// See https://llvm.org/LICENSE.txt for license information.
5+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6+
//
7+
//===----------------------------------------------------------------------===//
8+
9+
#pragma once
10+
11+
namespace sycl {
12+
inline namespace _V1 {
13+
namespace ext::oneapi::experimental {
14+
15+
enum class event_mode_enum { none, low_power };
16+
17+
struct event_mode
18+
: detail::run_time_property_key<event_mode, detail::PropKind::EventMode> {
19+
event_mode(event_mode_enum mode) : value(mode) {}
20+
21+
event_mode_enum value;
22+
};
23+
24+
using event_mode_key = event_mode;
25+
26+
inline bool operator==(const event_mode &lhs, const event_mode &rhs) {
27+
return lhs.value == rhs.value;
28+
}
29+
inline bool operator!=(const event_mode &lhs, const event_mode &rhs) {
30+
return !(lhs == rhs);
31+
}
32+
33+
} // namespace ext::oneapi::experimental
34+
} // namespace _V1
35+
} // namespace sycl

‎sycl/include/sycl/ext/oneapi/properties/property.hpp‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -221,8 +221,9 @@ enum PropKind : uint32_t {
221221
Prefetch = 76,
222222
Deterministic = 77,
223223
InitializeToIdentity = 78,
224+
EventMode = 79,
224225
// PropKindSize must always be the last value.
225-
PropKindSize = 79,
226+
PropKindSize = 80,
226227
};
227228

228229
template <typename PropertyT> struct PropertyToKind {

‎sycl/include/sycl/queue.hpp‎

Lines changed: 61 additions & 22 deletions
Original file line numberDiff line numberDiff line change
@@ -31,15 +31,16 @@
3131
#include <sycl/exception_list.hpp> // for defaultAsyncHa...
3232
#include <sycl/ext/oneapi/device_global/device_global.hpp> // for device_global
3333
#include <sycl/ext/oneapi/device_global/properties.hpp> // for device_image_s...
34-
#include <sycl/ext/oneapi/experimental/graph.hpp> // for command_graph...
35-
#include <sycl/ext/oneapi/properties/properties.hpp> // for empty_properti...
36-
#include <sycl/handler.hpp> // for handler, isDev...
37-
#include <sycl/id.hpp> // for id
38-
#include <sycl/kernel.hpp> // for auto_name
39-
#include <sycl/kernel_handler.hpp> // for kernel_handler
40-
#include <sycl/nd_range.hpp> // for nd_range
41-
#include <sycl/property_list.hpp> // for property_list
42-
#include <sycl/range.hpp> // for range
34+
#include <sycl/ext/oneapi/experimental/event_mode_property.hpp>
35+
#include <sycl/ext/oneapi/experimental/graph.hpp> // for command_graph...
36+
#include <sycl/ext/oneapi/properties/properties.hpp> // for empty_properti...
37+
#include <sycl/handler.hpp> // for handler, isDev...
38+
#include <sycl/id.hpp> // for id
39+
#include <sycl/kernel.hpp> // for auto_name
40+
#include <sycl/kernel_handler.hpp> // for kernel_handler
41+
#include <sycl/nd_range.hpp> // for nd_range
42+
#include <sycl/property_list.hpp> // for property_list
43+
#include <sycl/range.hpp> // for range
4344

4445
#include <cstddef> // for size_t
4546
#include <functional> // for function
@@ -103,6 +104,9 @@ class __SYCL_EXPORT SubmissionInfo {
103104
std::shared_ptr<detail::queue_impl> &SecondaryQueue();
104105
const std::shared_ptr<detail::queue_impl> &SecondaryQueue() const;
105106

107+
ext::oneapi::experimental::event_mode_enum &EventMode();
108+
const ext::oneapi::experimental::event_mode_enum &EventMode() const;
109+
106110
private:
107111
std::shared_ptr<SubmissionInfoImpl> impl = nullptr;
108112
};
@@ -115,9 +119,14 @@ enum class queue_state { executing, recording };
115119
struct image_descriptor;
116120

117121
namespace detail {
118-
template <typename CommandGroupFunc>
119-
void submit_impl(queue &Q, CommandGroupFunc &&CGF,
122+
template <typename CommandGroupFunc, typename PropertiesT>
123+
void submit_impl(queue &Q, PropertiesT Props, CommandGroupFunc &&CGF,
120124
const sycl::detail::code_location &CodeLoc);
125+
126+
template <typename CommandGroupFunc, typename PropertiesT>
127+
event submit_with_event_impl(queue &Q, PropertiesT Props,
128+
CommandGroupFunc &&CGF,
129+
const sycl::detail::code_location &CodeLoc);
121130
} // namespace detail
122131
} // namespace ext::oneapi::experimental
123132

@@ -365,7 +374,9 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
365374
std::enable_if_t<std::is_invocable_r_v<void, T, handler &>, event> submit(
366375
T CGF,
367376
const detail::code_location &CodeLoc = detail::code_location::current()) {
368-
return submit_with_event(CGF, /*SecondaryQueuePtr=*/nullptr, CodeLoc);
377+
return submit_with_event(
378+
sycl::ext::oneapi::experimental::empty_properties_t{}, CGF,
379+
/*SecondaryQueuePtr=*/nullptr, CodeLoc);
369380
}
370381

371382
/// Submits a command group function object to the queue, in order to be
@@ -383,7 +394,9 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
383394
std::enable_if_t<std::is_invocable_r_v<void, T, handler &>, event> submit(
384395
T CGF, queue &SecondaryQueue,
385396
const detail::code_location &CodeLoc = detail::code_location::current()) {
386-
return submit_with_event(CGF, &SecondaryQueue, CodeLoc);
397+
return submit_with_event(
398+
sycl::ext::oneapi::experimental::empty_properties_t{}, CGF,
399+
&SecondaryQueue, CodeLoc);
387400
}
388401

389402
/// Prevents any commands submitted afterward to this queue from executing
@@ -2746,11 +2759,28 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
27462759
const detail::code_location &);
27472760
#endif
27482761

2749-
template <typename CommandGroupFunc>
2762+
template <typename CommandGroupFunc, typename PropertiesT>
27502763
friend void ext::oneapi::experimental::detail::submit_impl(
2751-
queue &Q, CommandGroupFunc &&CGF,
2764+
queue &Q, PropertiesT Props, CommandGroupFunc &&CGF,
27522765
const sycl::detail::code_location &CodeLoc);
27532766

2767+
template <typename CommandGroupFunc, typename PropertiesT>
2768+
friend event ext::oneapi::experimental::detail::submit_with_event_impl(
2769+
queue &Q, PropertiesT Props, CommandGroupFunc &&CGF,
2770+
const sycl::detail::code_location &CodeLoc);
2771+
2772+
template <typename PropertiesT>
2773+
void ProcessSubmitProperties(PropertiesT Props, detail::SubmissionInfo &SI) {
2774+
if constexpr (Props.template has_property<
2775+
ext::oneapi::experimental::event_mode_key>()) {
2776+
ext::oneapi::experimental::event_mode EventModeProp =
2777+
Props.template get_property<ext::oneapi::experimental::event_mode>();
2778+
if (EventModeProp.value !=
2779+
ext::oneapi::experimental::event_mode_enum::none)
2780+
SI.EventMode() = EventModeProp.value;
2781+
}
2782+
}
2783+
27542784
#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
27552785
/// TODO: Unused. Remove these when ABI-break window is open.
27562786
event submit_impl(std::function<void(handler &)> CGH,
@@ -2799,16 +2829,18 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
27992829
/// Submits a command group function object to the queue, in order to be
28002830
/// scheduled for execution on the device.
28012831
///
2832+
/// \param Props is a property list with submission properties.
28022833
/// \param CGF is a function object containing command group.
28032834
/// \param CodeLoc is the code location of the submit call (default argument)
28042835
/// \return a SYCL event object for the submitted command group.
2805-
template <typename T>
2836+
template <typename T, typename PropertiesT>
28062837
std::enable_if_t<std::is_invocable_r_v<void, T, handler &>, event>
28072838
submit_with_event(
2808-
T CGF, queue *SecondaryQueuePtr,
2839+
PropertiesT Props, T CGF, queue *SecondaryQueuePtr,
28092840
const detail::code_location &CodeLoc = detail::code_location::current()) {
28102841
detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
28112842
detail::SubmissionInfo SI{};
2843+
ProcessSubmitProperties(Props, SI);
28122844
if (SecondaryQueuePtr)
28132845
SI.SecondaryQueue() = detail::getSyclObjImpl(*SecondaryQueuePtr);
28142846
#if __SYCL_USE_FALLBACK_ASSERT
@@ -2833,18 +2865,21 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
28332865
/// Submits a command group function object to the queue, in order to be
28342866
/// scheduled for execution on the device.
28352867
///
2868+
/// \param Props is a property list with submission properties.
28362869
/// \param CGF is a function object containing command group.
28372870
/// \param CodeLoc is the code location of the submit call (default argument)
2838-
template <typename T>
2871+
template <typename T, typename PropertiesT>
28392872
std::enable_if_t<std::is_invocable_r_v<void, T, handler &>, void>
2840-
submit_without_event(T CGF, const detail::code_location &CodeLoc) {
2873+
submit_without_event(PropertiesT Props, T CGF,
2874+
const detail::code_location &CodeLoc) {
28412875
#if __SYCL_USE_FALLBACK_ASSERT
28422876
// If post-processing is needed, fall back to the regular submit.
28432877
// TODO: Revisit whether we can avoid this.
2844-
submit_with_event(CGF, nullptr, CodeLoc);
2878+
submit_with_event(Props, CGF, nullptr, CodeLoc);
28452879
#else
28462880
detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
28472881
detail::SubmissionInfo SI{};
2882+
ProcessSubmitProperties(Props, SI);
28482883
submit_without_event_impl(CGF, SI, TlsCodeLocCapture.query(),
28492884
TlsCodeLocCapture.isToplevel());
28502885
#endif // __SYCL_USE_FALLBACK_ASSERT
@@ -3071,8 +3106,12 @@ event submitAssertCapture(queue &Self, event &Event, queue *SecondaryQueue,
30713106
});
30723107
};
30733108

3074-
CopierEv = Self.submit_with_event(CopierCGF, SecondaryQueue, CodeLoc);
3075-
CheckerEv = Self.submit_with_event(CheckerCGF, SecondaryQueue, CodeLoc);
3109+
CopierEv = Self.submit_with_event(
3110+
sycl::ext::oneapi::experimental::empty_properties_t{}, CopierCGF,
3111+
SecondaryQueue, CodeLoc);
3112+
CheckerEv = Self.submit_with_event(
3113+
sycl::ext::oneapi::experimental::empty_properties_t{}, CheckerCGF,
3114+
SecondaryQueue, CodeLoc);
30763115

30773116
return CheckerEv;
30783117
}

‎sycl/source/detail/cg.hpp‎

Lines changed: 8 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -15,8 +15,9 @@
1515
#include <sycl/detail/ur.hpp> // for ur_rect_region_t, ur_rect_offset_t
1616
#include <sycl/event.hpp> // for event_impl
1717
#include <sycl/exception_list.hpp> // for queue_impl
18-
#include <sycl/kernel.hpp> // for kernel_impl
19-
#include <sycl/kernel_bundle.hpp> // for kernel_bundle_impl
18+
#include <sycl/ext/oneapi/experimental/event_mode_property.hpp>
19+
#include <sycl/kernel.hpp> // for kernel_impl
20+
#include <sycl/kernel_bundle.hpp> // for kernel_bundle_impl
2021

2122
#include <assert.h> // for assert
2223
#include <memory> // for shared_ptr, unique_ptr
@@ -423,12 +424,16 @@ class CGAdviseUSM : public CG {
423424
class CGBarrier : public CG {
424425
public:
425426
std::vector<detail::EventImplPtr> MEventsWaitWithBarrier;
427+
ext::oneapi::experimental::event_mode_enum MEventMode =
428+
ext::oneapi::experimental::event_mode_enum::none;
426429

427430
CGBarrier(std::vector<detail::EventImplPtr> EventsWaitWithBarrier,
431+
ext::oneapi::experimental::event_mode_enum EventMode,
428432
CG::StorageInitHelper CGData, CGType Type,
429433
detail::code_location loc = {})
430434
: CG(Type, std::move(CGData), std::move(loc)),
431-
MEventsWaitWithBarrier(std::move(EventsWaitWithBarrier)) {}
435+
MEventsWaitWithBarrier(std::move(EventsWaitWithBarrier)),
436+
MEventMode(EventMode) {}
432437
};
433438

434439
class CGProfilingTag : public CG {

‎sycl/source/detail/handler_impl.hpp‎

Lines changed: 6 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -195,7 +195,12 @@ class handler_impl {
195195
bool MIsTopCodeLoc = true;
196196

197197
/// List of work group memory objects associated with this handler
198-
std::vector<std::shared_ptr<detail::work_group_memory_impl>> MWorkGroupMemoryObjects;
198+
std::vector<std::shared_ptr<detail::work_group_memory_impl>>
199+
MWorkGroupMemoryObjects;
200+
201+
/// Potential event mode for the result event of the command.
202+
ext::oneapi::experimental::event_mode_enum MEventMode =
203+
ext::oneapi::experimental::event_mode_enum::none;
199204
};
200205

201206
} // namespace detail

‎sycl/source/detail/queue_impl.cpp‎

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -357,6 +357,7 @@ event queue_impl::submit_impl(const std::function<void(handler &)> &CGF,
357357
bool IsTopCodeLoc,
358358
const SubmissionInfo &SubmitInfo) {
359359
handler Handler(Self, PrimaryQueue, SecondaryQueue, CallerNeedsEvent);
360+
auto HandlerImpl = detail::getSyclObjImpl(Handler);
360361
Handler.saveCodeLoc(Loc, IsTopCodeLoc);
361362

362363
{
@@ -367,13 +368,15 @@ event queue_impl::submit_impl(const std::function<void(handler &)> &CGF,
367368
// Scheduler will later omit events, that are not required to execute tasks.
368369
// Host and interop tasks, however, are not submitted to low-level runtimes
369370
// and require separate dependency management.
370-
const CGType Type = detail::getSyclObjImpl(Handler)->MCGType;
371+
const CGType Type = HandlerImpl->MCGType;
371372
event Event = detail::createSyclObjFromImpl<event>(
372373
std::make_shared<detail::event_impl>());
373374
std::vector<StreamImplPtr> Streams;
374375
if (Type == CGType::Kernel)
375376
Streams = std::move(Handler.MStreamStorage);
376377

378+
HandlerImpl->MEventMode = SubmitInfo.EventMode();
379+
377380
if (SubmitInfo.PostProcessorFunc()) {
378381
auto &PostProcess = *SubmitInfo.PostProcessorFunc();
379382

‎sycl/source/detail/queue_impl.hpp‎

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -71,6 +71,8 @@ enum QueueOrder { Ordered, OOO };
7171
struct SubmissionInfoImpl {
7272
optional<detail::SubmitPostProcessF> MPostProcessorFunc = std::nullopt;
7373
std::shared_ptr<detail::queue_impl> MSecondaryQueue = nullptr;
74+
ext::oneapi::experimental::event_mode_enum MEventMode =
75+
ext::oneapi::experimental::event_mode_enum::none;
7476
};
7577

7678
class queue_impl {

0 commit comments

Comments
 (0)