Skip to content

Commit 6309d6d

Browse files
mateuszpnmateuszpn
authored
[SYCL][L0 v2] Use counter-based events (#20788)
Use counter-based events in level-zero v2 adapter --------- Signed-off-by: Mateusz P. Nowak <mateusz.p.nowak@intel.com>
1 parent 5604623 commit 6309d6d

File tree

11 files changed

+123
-40
lines changed

11 files changed

+123
-40
lines changed

‎sycl/test-e2e/Adapters/level_zero/event-leak.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,8 @@
11
// REQUIRES: level_zero, level_zero_dev_kit
22
//
3+
// UNSUPPORTED: windows && level_zero_v2_adapter
4+
// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20852
5+
//
36
// RUN: %{build} %level_zero_options -o %t.out
47
// RUN: %{l0_leak_check} %{run} %t.out wait 2>&1 | FileCheck %s
58
// RUN: %{l0_leak_check} %{run} %t.out nowait 2>&1 | FileCheck %s

‎sycl/test-e2e/DeprecatedFeatures/DiscardEvents/discard_events_l0_leak.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,8 @@
11
// REQUIRES: level_zero
22
//
3+
// UNSUPPORTED: windows && level_zero_v2_adapter
4+
// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20852
5+
//
36
// RUN: %{build} -o %t.out
47
//
58
// RUN: env SYCL_PI_LEVEL_ZERO_USE_IMMEDIATE_COMMANDLISTS=0 SYCL_PI_LEVEL_ZERO_BATCH_SIZE=4 ONEAPI_DEVICE_SELECTOR='level_zero:*' %{l0_leak_check} %{run} %t.out wait 2>&1 | FileCheck %s

‎sycl/test-e2e/ProfilingTag/profile_tag_leak.cpp‎

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,8 @@
11
// REQUIRES: level_zero
2-
2+
//
3+
// UNSUPPORTED: windows && level_zero_v2_adapter
4+
// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20852
5+
//
36
// RUN: %{build} -o %t.out
47
// RUN: %{l0_leak_check} %{run} %t.out 2>&1 | FileCheck %s --implicit-check-not=LEAK
58

‎sycl/test-e2e/Regression/reduction_resource_leak_dw.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,8 @@
11
// REQUIRES: level_zero
22
//
3+
// UNSUPPORTED: windows && level_zero_v2_adapter
4+
// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20852
5+
//
36
// RUN: %{build} -o %t.out
47
// RUN: %{l0_leak_check} %{run} %t.out 2>&1 | FileCheck %s
58
//

‎sycl/test-e2e/USM/usm_leak_check.cpp‎

Lines changed: 5 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,7 @@
11
// REQUIRES: level_zero
2+
//
3+
// UNSUPPORTED: windows && level_zero_v2_adapter
4+
// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20852
25

36
// RUN: %{build} -Wno-error=deprecated-declarations -o %t.out
47

@@ -36,10 +39,8 @@ template <typename T, size_t N> void sycl_buffer(queue &Q) {
3639
accessor accB{bufferB, cgh, read_only};
3740
accessor accC{bufferC, cgh, write_only};
3841

39-
cgh.parallel_for<class K<T>>(numElems,
40-
[=](id<1> wiID) {
41-
accC[wiID] = accA[wiID] + accB[wiID];
42-
});
42+
cgh.parallel_for<class K<T>>(
43+
numElems, [=](id<1> wiID) { accC[wiID] = accA[wiID] + accB[wiID]; });
4344
});
4445
}
4546

‎unified-runtime/source/adapters/level_zero/v2/context.cpp‎

Lines changed: 18 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -80,23 +80,26 @@ ur_context_handle_t_::ur_context_handle_t_(ze_context_handle_t hContext,
8080
phDevices[0]->Platform->ZeMutableCmdListExt.Supported}),
8181
eventPoolCacheImmediate(
8282
this, phDevices[0]->Platform->getNumDevices(),
83-
[context = this](DeviceId /* deviceId*/, v2::event_flags_t flags)
84-
-> std::unique_ptr<v2::event_provider> {
83+
[context = this, platform = phDevices[0]->Platform](
84+
DeviceId deviceId,
85+
v2::event_flags_t flags) -> std::unique_ptr<v2::event_provider> {
86+
auto device = platform->getDeviceById(deviceId);
87+
88+
// TODO: just use per-context id?
89+
return v2::createProvider(platform, context, v2::QUEUE_IMMEDIATE,
90+
device, flags);
91+
}),
92+
eventPoolCacheRegular(
93+
this, phDevices[0]->Platform->getNumDevices(),
94+
[context = this, platform = phDevices[0]->Platform](
95+
DeviceId deviceId,
96+
v2::event_flags_t flags) -> std::unique_ptr<v2::event_provider> {
97+
auto device = platform->getDeviceById(deviceId);
98+
8599
// TODO: just use per-context id?
86-
return std::make_unique<v2::provider_normal>(
87-
context, v2::QUEUE_IMMEDIATE, flags);
100+
return v2::createProvider(platform, context, v2::QUEUE_REGULAR,
101+
device, flags);
88102
}),
89-
eventPoolCacheRegular(this, phDevices[0]->Platform->getNumDevices(),
90-
[context = this, platform = phDevices[0]->Platform](
91-
DeviceId deviceId, v2::event_flags_t flags)
92-
-> std::unique_ptr<v2::event_provider> {
93-
std::ignore = deviceId;
94-
std::ignore = platform;
95-
96-
// TODO: just use per-context id?
97-
return std::make_unique<v2::provider_normal>(
98-
context, v2::QUEUE_REGULAR, flags);
99-
}),
100103
nativeEventsPool(this, std::make_unique<v2::provider_normal>(
101104
this, v2::QUEUE_IMMEDIATE,
102105
v2::EVENT_FLAGS_PROFILING_ENABLED)),

‎unified-runtime/source/adapters/level_zero/v2/event_provider.hpp‎

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -28,6 +28,11 @@ enum event_flag_t {
2828
};
2929
static constexpr size_t EVENT_FLAGS_USED_BITS = 2;
3030

31+
enum queue_type {
32+
QUEUE_REGULAR,
33+
QUEUE_IMMEDIATE,
34+
};
35+
3136
class event_provider;
3237

3338
namespace raii {

‎unified-runtime/source/adapters/level_zero/v2/event_provider_counter.cpp‎

Lines changed: 63 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -13,6 +13,7 @@
1313
#include "context.hpp"
1414
#include "event_provider.hpp"
1515
#include "event_provider_counter.hpp"
16+
#include "event_provider_normal.hpp"
1617
#include "loader/ze_loader.h"
1718

1819
#include "../device.hpp"
@@ -22,10 +23,17 @@ namespace v2 {
2223

2324
provider_counter::provider_counter(ur_platform_handle_t platform,
2425
ur_context_handle_t context,
25-
ur_device_handle_t device) {
26+
queue_type queueType,
27+
ur_device_handle_t device,
28+
event_flags_t flags)
29+
: queueType(queueType), flags(flags) {
30+
assert(flags & EVENT_FLAGS_COUNTER);
31+
32+
// Try to get the counter-based event extension function
2633
ZE2UR_CALL_THROWS(zeDriverGetExtensionFunctionAddress,
27-
(platform->ZeDriver, "zexCounterBasedEventCreate",
34+
(platform->ZeDriver, "zexCounterBasedEventCreate2",
2835
(void **)&this->eventCreateFunc));
36+
2937
ZE2UR_CALL_THROWS(zelLoaderTranslateHandle,
3038
(ZEL_HANDLE_CONTEXT, context->getZeHandle(),
3139
(void **)&translatedContext));
@@ -34,17 +42,41 @@ provider_counter::provider_counter(ur_platform_handle_t platform,
3442
(ZEL_HANDLE_DEVICE, device->ZeDevice, (void **)&translatedDevice));
3543
}
3644

45+
static zex_counter_based_event_exp_flags_t createZeFlags(queue_type queueType,
46+
event_flags_t flags) {
47+
zex_counter_based_event_exp_flags_t zeFlags =
48+
ZEX_COUNTER_BASED_EVENT_FLAG_HOST_VISIBLE;
49+
if (flags & EVENT_FLAGS_PROFILING_ENABLED) {
50+
zeFlags |= ZEX_COUNTER_BASED_EVENT_FLAG_KERNEL_TIMESTAMP;
51+
}
52+
53+
if (queueType == QUEUE_IMMEDIATE) {
54+
zeFlags |= ZEX_COUNTER_BASED_EVENT_FLAG_IMMEDIATE;
55+
} else {
56+
zeFlags |= ZEX_COUNTER_BASED_EVENT_FLAG_NON_IMMEDIATE;
57+
}
58+
59+
return zeFlags;
60+
}
61+
3762
raii::cache_borrowed_event provider_counter::allocate() {
3863
if (freelist.empty()) {
39-
ZeStruct<ze_event_desc_t> desc;
40-
desc.index = 0;
41-
desc.signal = ZE_EVENT_SCOPE_FLAG_HOST;
42-
desc.wait = 0;
64+
zex_counter_based_event_desc_t desc = {};
65+
desc.stype = ZEX_STRUCTURE_COUNTER_BASED_EVENT_DESC;
66+
desc.flags = createZeFlags(queueType, flags);
67+
desc.signalScope = ZE_EVENT_SCOPE_FLAG_HOST;
68+
69+
uint32_t equivalentFlags = ZE_EVENT_POOL_FLAG_HOST_VISIBLE;
70+
if (flags & EVENT_FLAGS_PROFILING_ENABLED) {
71+
equivalentFlags |= ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP;
72+
}
73+
UR_LOG(DEBUG, "ze_event_pool_desc_t flags set to: {}", equivalentFlags);
74+
4375
ze_event_handle_t handle;
4476

4577
// TODO: allocate host and device buffers to use here
46-
ZE2UR_CALL_THROWS(eventCreateFunc, (translatedContext, translatedDevice,
47-
nullptr, nullptr, 0, &desc, &handle));
78+
ZE2UR_CALL_THROWS(eventCreateFunc,
79+
(translatedContext, translatedDevice, &desc, &handle));
4880

4981
freelist.emplace_back(handle);
5082
}
@@ -57,8 +89,29 @@ raii::cache_borrowed_event provider_counter::allocate() {
5789
[this](ze_event_handle_t handle) { freelist.push_back(handle); });
5890
}
5991

60-
event_flags_t provider_counter::eventFlags() const {
61-
return EVENT_FLAGS_COUNTER;
92+
event_flags_t provider_counter::eventFlags() const { return flags; }
93+
94+
std::unique_ptr<event_provider> createProvider(ur_platform_handle_t platform,
95+
ur_context_handle_t context,
96+
queue_type queueType,
97+
ur_device_handle_t device,
98+
event_flags_t flags) {
99+
// Only try counter-based events if the flag is set
100+
if (flags & EVENT_FLAGS_COUNTER) {
101+
// Try to create a counter-based event provider first
102+
try {
103+
return std::make_unique<provider_counter>(platform, context, queueType,
104+
device, flags);
105+
} catch (...) {
106+
// If the new counter-based API (zexCounterBasedEventCreate2) is not
107+
// available, fall back to normal provider which support counter-based
108+
// events using the old API
109+
return std::make_unique<provider_normal>(context, queueType, flags);
110+
}
111+
}
112+
113+
// Counter-based events not requested, use normal events
114+
return std::make_unique<provider_normal>(context, queueType, flags);
62115
}
63116

64117
} // namespace v2

‎unified-runtime/source/adapters/level_zero/v2/event_provider_counter.hpp‎

Lines changed: 16 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -25,22 +25,27 @@
2525

2626
#include "../device.hpp"
2727

28+
#include <level_zero/driver_experimental/zex_event.h>
29+
#include <level_zero/ze_intel_gpu.h>
30+
2831
namespace v2 {
2932

3033
typedef ze_result_t (*zexCounterBasedEventCreate)(
3134
ze_context_handle_t hContext, ze_device_handle_t hDevice,
32-
uint64_t *deviceAddress, uint64_t *hostAddress, uint64_t completionValue,
33-
const ze_event_desc_t *desc, ze_event_handle_t *phEvent);
35+
const zex_counter_based_event_desc_t *desc, ze_event_handle_t *phEvent);
3436

3537
class provider_counter : public event_provider {
3638
public:
3739
provider_counter(ur_platform_handle_t platform, ur_context_handle_t,
38-
ur_device_handle_t);
40+
queue_type, ur_device_handle_t, event_flags_t);
3941

4042
raii::cache_borrowed_event allocate() override;
4143
event_flags_t eventFlags() const override;
4244

4345
private:
46+
queue_type queueType;
47+
event_flags_t flags;
48+
4449
ze_context_handle_t translatedContext;
4550
ze_device_handle_t translatedDevice;
4651

@@ -49,4 +54,12 @@ class provider_counter : public event_provider {
4954
std::vector<raii::ze_event_handle_t> freelist;
5055
};
5156

57+
// Factory function that creates a counter-based provider with fallback to
58+
// normal provider
59+
std::unique_ptr<event_provider> createProvider(ur_platform_handle_t platform,
60+
ur_context_handle_t context,
61+
queue_type queueType,
62+
ur_device_handle_t device,
63+
event_flags_t flags);
64+
5265
} // namespace v2

‎unified-runtime/source/adapters/level_zero/v2/event_provider_normal.hpp‎

Lines changed: 1 addition & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -21,17 +21,13 @@
2121

2222
#include "common.hpp"
2323
#include "event.hpp"
24+
#include "event_provider.hpp"
2425

2526
#include "../device.hpp"
2627
#include "../ur_interface_loader.hpp"
2728

2829
namespace v2 {
2930

30-
enum queue_type {
31-
QUEUE_REGULAR,
32-
QUEUE_IMMEDIATE,
33-
};
34-
3531
class provider_pool {
3632
public:
3733
provider_pool(ur_context_handle_t, queue_type, event_flags_t flags);

0 commit comments

Comments
 (0)