Skip to content

Commit f378976

Browse files
AlicjaLukaszewiczAlicjaLukaszewicz
authored
fix: calculate timeout based on device responsiveness (#234)
* fix: calculate timeout based on device responsiveness Related-To: VCV-13070 Signed-off-by: Alicja Lukaszewicz <[email protected]>
1 parent 32b3e68 commit f378976

File tree

5 files changed

+224
-5
lines changed

5 files changed

+224
-5
lines changed

‎conformance_tests/core/test_cmdlist/src/test_cmdlist_immediate.cpp

Lines changed: 69 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1469,6 +1469,74 @@ class zeImmediateCommandListHostSynchronizeTimeoutTests
14691469
cl = lzt::create_immediate_command_list(
14701470
context, device, cq_flags, ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS,
14711471
ZE_COMMAND_QUEUE_PRIORITY_NORMAL, 0, 0);
1472+
1473+
const uint64_t responsiveness = measure_device_responsiveness();
1474+
const double min_ratio = 0.02;
1475+
if (responsiveness > static_cast<uint64_t>(min_ratio * timeout)) {
1476+
timeout = static_cast<uint64_t>(responsiveness / min_ratio);
1477+
LOG_INFO << "Device responsiveness: " << responsiveness
1478+
<< " ns, setting timeout to: " << timeout << " ns";
1479+
}
1480+
}
1481+
1482+
uint64_t measure_device_responsiveness() {
1483+
const ze_context_handle_t context =
1484+
lzt::create_context(lzt::get_default_driver());
1485+
const ze_device_handle_t device =
1486+
lzt::get_default_device(lzt::get_default_driver());
1487+
ze_command_list_handle_t cmd_list = lzt::create_immediate_command_list(
1488+
device, 0, ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS,
1489+
ZE_COMMAND_QUEUE_PRIORITY_NORMAL, 0);
1490+
1491+
ze_module_handle_t module =
1492+
lzt::create_module(context, device, "cmdlist_add.spv",
1493+
ZE_MODULE_FORMAT_IL_SPIRV, nullptr, nullptr);
1494+
ze_kernel_handle_t kernel =
1495+
lzt::create_function(module, "cmdlist_add_constant");
1496+
1497+
size_t size = 10000;
1498+
size_t buff_size = size * sizeof(int);
1499+
int *buf_hst =
1500+
static_cast<int *>(lzt::allocate_host_memory(buff_size, 1, context));
1501+
int *buf_dev = static_cast<int *>(
1502+
lzt::allocate_device_memory(buff_size, 1, 0, 0, device, context));
1503+
std::fill_n(buf_hst, size, 0);
1504+
1505+
const int add_value = 7;
1506+
lzt::set_group_size(kernel, 1, 1, 1);
1507+
ze_group_count_t args = {static_cast<uint32_t>(size), 1, 1};
1508+
lzt::set_argument_value(kernel, 0, sizeof(buf_dev), &buf_dev);
1509+
lzt::set_argument_value(kernel, 1, sizeof(add_value), &add_value);
1510+
1511+
lzt::append_memory_copy(cmd_list, buf_dev, buf_hst, buff_size, nullptr, 0,
1512+
nullptr);
1513+
1514+
const auto t0 = std::chrono::steady_clock::now();
1515+
for (size_t i = 0; i < 2; ++i) {
1516+
lzt::append_launch_function(cmd_list, kernel, &args, nullptr, 0, nullptr);
1517+
}
1518+
const auto t1 = std::chrono::steady_clock::now();
1519+
1520+
lzt::append_memory_copy(cmd_list, buf_hst, buf_dev, buff_size, nullptr, 0,
1521+
nullptr);
1522+
lzt::synchronize_command_list_host(cmd_list,
1523+
std::numeric_limits<uint64_t>::max());
1524+
1525+
// Verify
1526+
for (size_t i = 0; i < size; ++i) {
1527+
EXPECT_EQ(buf_hst[i], 2 * add_value);
1528+
}
1529+
1530+
// Cleanup
1531+
lzt::destroy_function(kernel);
1532+
lzt::destroy_module(module);
1533+
lzt::free_memory(context, buf_dev);
1534+
lzt::free_memory(context, buf_hst);
1535+
lzt::destroy_command_list(cmd_list);
1536+
lzt::destroy_context(context);
1537+
1538+
return std::chrono::duration_cast<std::chrono::nanoseconds>(t1 - t0)
1539+
.count();
14721540
}
14731541

14741542
void TearDown() override {
@@ -1477,7 +1545,7 @@ class zeImmediateCommandListHostSynchronizeTimeoutTests
14771545
lzt::destroy_command_list(cl);
14781546
}
14791547

1480-
const uint64_t timeout = 5000000;
1548+
uint64_t timeout = 5000000;
14811549
ze_event_pool_handle_t ep = nullptr;
14821550
ze_event_handle_t ev = nullptr;
14831551
ze_command_list_handle_t cl = nullptr;

‎conformance_tests/core/test_cmdqueue/kernels/cmdqueue_add.cl

Lines changed: 13 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,13 @@
1+
/*
2+
*
3+
* Copyright (C) 2025 Intel Corporation
4+
*
5+
* SPDX-License-Identifier: MIT
6+
*
7+
*/
8+
9+
kernel void cmdqueue_add_constant(global int *values_in, global int *values_out, int addval) {
10+
11+
const int xid = get_global_id(0);
12+
values_out[xid] = values_in[xid] + addval;
13+
}

‎conformance_tests/core/test_cmdqueue/kernels/cmdqueue_add.spv

1.3 KB
Binary file not shown.

‎conformance_tests/core/test_cmdqueue/src/test_cmdqueue.cpp

Lines changed: 71 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
/*
22
*
3-
* Copyright (C) 2019-2024 Intel Corporation
3+
* Copyright (C) 2019-2025 Intel Corporation
44
*
55
* SPDX-License-Identifier: MIT
66
*
@@ -1049,6 +1049,75 @@ class zeCommandQueueSynchronizeTimeoutTests
10491049
ZE_COMMAND_QUEUE_PRIORITY_NORMAL, 0, 0);
10501050

10511051
cl = lzt::create_command_list(context, device, cl_flags, 0);
1052+
1053+
const uint64_t responsiveness = measure_device_responsiveness();
1054+
const double min_ratio = 0.02;
1055+
if (responsiveness > static_cast<uint64_t>(min_ratio * timeout)) {
1056+
timeout = static_cast<uint64_t>(responsiveness / min_ratio);
1057+
LOG_INFO << "Device responsiveness: " << responsiveness
1058+
<< " ns, setting timeout to: " << timeout << " ns";
1059+
}
1060+
}
1061+
1062+
uint64_t measure_device_responsiveness() {
1063+
const ze_context_handle_t context =
1064+
lzt::create_context(lzt::get_default_driver());
1065+
const ze_device_handle_t device =
1066+
lzt::get_default_device(lzt::get_default_driver());
1067+
ze_command_list_handle_t cmd_list = lzt::create_immediate_command_list(
1068+
device, 0, ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS,
1069+
ZE_COMMAND_QUEUE_PRIORITY_NORMAL, 0);
1070+
1071+
ze_module_handle_t module =
1072+
lzt::create_module(context, device, "cmdqueue_add.spv",
1073+
ZE_MODULE_FORMAT_IL_SPIRV, nullptr, nullptr);
1074+
ze_kernel_handle_t kernel =
1075+
lzt::create_function(module, "cmdqueue_add_constant");
1076+
1077+
size_t size = 10000;
1078+
size_t buff_size = size * sizeof(int);
1079+
int *buf_hst =
1080+
static_cast<int *>(lzt::allocate_host_memory(buff_size, 1, context));
1081+
int *buf_dev = static_cast<int *>(
1082+
lzt::allocate_device_memory(buff_size, 1, 0, 0, device, context));
1083+
std::fill_n(buf_hst, size, 0);
1084+
1085+
const int add_value = 7;
1086+
lzt::set_group_size(kernel, 1, 1, 1);
1087+
ze_group_count_t args = {static_cast<uint32_t>(size), 1, 1};
1088+
lzt::set_argument_value(kernel, 0, sizeof(buf_dev), &buf_dev);
1089+
lzt::set_argument_value(kernel, 1, sizeof(buf_dev), &buf_dev);
1090+
lzt::set_argument_value(kernel, 2, sizeof(add_value), &add_value);
1091+
1092+
lzt::append_memory_copy(cmd_list, buf_dev, buf_hst, buff_size, nullptr, 0,
1093+
nullptr);
1094+
1095+
const auto t0 = std::chrono::steady_clock::now();
1096+
for (size_t i = 0; i < 2; ++i) {
1097+
lzt::append_launch_function(cmd_list, kernel, &args, nullptr, 0, nullptr);
1098+
}
1099+
const auto t1 = std::chrono::steady_clock::now();
1100+
1101+
lzt::append_memory_copy(cmd_list, buf_hst, buf_dev, buff_size, nullptr, 0,
1102+
nullptr);
1103+
lzt::synchronize_command_list_host(cmd_list,
1104+
std::numeric_limits<uint64_t>::max());
1105+
1106+
// Verify
1107+
for (size_t i = 0; i < size; ++i) {
1108+
EXPECT_EQ(buf_hst[i], 2 * add_value);
1109+
}
1110+
1111+
// Cleanup
1112+
lzt::destroy_function(kernel);
1113+
lzt::destroy_module(module);
1114+
lzt::free_memory(context, buf_dev);
1115+
lzt::free_memory(context, buf_hst);
1116+
lzt::destroy_command_list(cmd_list);
1117+
lzt::destroy_context(context);
1118+
1119+
return std::chrono::duration_cast<std::chrono::nanoseconds>(t1 - t0)
1120+
.count();
10521121
}
10531122

10541123
void TearDown() override {
@@ -1058,7 +1127,7 @@ class zeCommandQueueSynchronizeTimeoutTests
10581127
lzt::destroy_command_queue(cq);
10591128
}
10601129

1061-
const uint64_t timeout = 5000000;
1130+
uint64_t timeout = 5000000;
10621131
ze_event_pool_handle_t ep = nullptr;
10631132
ze_event_handle_t ev = nullptr;
10641133
ze_command_queue_handle_t cq = nullptr;

‎conformance_tests/core/test_event/src/test_event.cpp

Lines changed: 71 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
/*
22
*
3-
* Copyright (C) 2019-2023 Intel Corporation
3+
* Copyright (C) 2019-2025 Intel Corporation
44
*
55
* SPDX-License-Identifier: MIT
66
*
@@ -733,14 +733,83 @@ class zeEventHostSynchronizeTimeoutTests
733733
ev_desc.signal = ZE_EVENT_SCOPE_FLAG_HOST;
734734
ev_desc.wait = ZE_EVENT_SCOPE_FLAG_HOST;
735735
ev = lzt::create_event(ep, ev_desc);
736+
737+
const uint64_t responsiveness = measure_device_responsiveness();
738+
const double min_ratio = 0.02;
739+
if (responsiveness > static_cast<uint64_t>(min_ratio * timeout)) {
740+
timeout = static_cast<uint64_t>(responsiveness / min_ratio);
741+
LOG_INFO << "Device responsiveness: " << responsiveness
742+
<< " ns, setting timeout to: " << timeout << " ns";
743+
}
744+
}
745+
746+
uint64_t measure_device_responsiveness() {
747+
const ze_context_handle_t context =
748+
lzt::create_context(lzt::get_default_driver());
749+
const ze_device_handle_t device =
750+
lzt::get_default_device(lzt::get_default_driver());
751+
ze_command_list_handle_t cmd_list = lzt::create_immediate_command_list(
752+
device, 0, ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS,
753+
ZE_COMMAND_QUEUE_PRIORITY_NORMAL, 0);
754+
755+
ze_module_handle_t module =
756+
lzt::create_module(context, device, "profile_add.spv",
757+
ZE_MODULE_FORMAT_IL_SPIRV, nullptr, nullptr);
758+
ze_kernel_handle_t kernel =
759+
lzt::create_function(module, "profile_add_constant");
760+
761+
size_t size = 10000;
762+
size_t buff_size = size * sizeof(int);
763+
int *buf_hst =
764+
static_cast<int *>(lzt::allocate_host_memory(buff_size, 1, context));
765+
int *buf_dev = static_cast<int *>(
766+
lzt::allocate_device_memory(buff_size, 1, 0, 0, device, context));
767+
std::fill_n(buf_hst, size, 0);
768+
769+
const int add_value = 7;
770+
lzt::set_group_size(kernel, 1, 1, 1);
771+
ze_group_count_t args = {static_cast<uint32_t>(size), 1, 1};
772+
lzt::set_argument_value(kernel, 0, sizeof(buf_dev), &buf_dev);
773+
lzt::set_argument_value(kernel, 1, sizeof(buf_dev), &buf_dev);
774+
lzt::set_argument_value(kernel, 2, sizeof(add_value), &add_value);
775+
776+
lzt::append_memory_copy(cmd_list, buf_dev, buf_hst, buff_size, nullptr, 0,
777+
nullptr);
778+
779+
const auto t0 = std::chrono::steady_clock::now();
780+
for (size_t i = 0; i < 2; ++i) {
781+
lzt::append_launch_function(cmd_list, kernel, &args, nullptr, 0, nullptr);
782+
}
783+
const auto t1 = std::chrono::steady_clock::now();
784+
785+
lzt::append_memory_copy(cmd_list, buf_hst, buf_dev, buff_size, nullptr, 0,
786+
nullptr);
787+
lzt::synchronize_command_list_host(cmd_list,
788+
std::numeric_limits<uint64_t>::max());
789+
790+
// Verify
791+
for (size_t i = 0; i < size; ++i) {
792+
EXPECT_EQ(buf_hst[i], 2 * add_value);
793+
}
794+
795+
// Cleanup
796+
lzt::destroy_function(kernel);
797+
lzt::destroy_module(module);
798+
lzt::free_memory(context, buf_dev);
799+
lzt::free_memory(context, buf_hst);
800+
lzt::destroy_command_list(cmd_list);
801+
lzt::destroy_context(context);
802+
803+
return std::chrono::duration_cast<std::chrono::nanoseconds>(t1 - t0)
804+
.count();
736805
}
737806

738807
void TearDown() override {
739808
lzt::destroy_event(ev);
740809
lzt::destroy_event_pool(ep);
741810
}
742811

743-
const uint64_t timeout = 5000000;
812+
uint64_t timeout = 5000000;
744813
ze_event_pool_handle_t ep = nullptr;
745814
ze_event_handle_t ev = nullptr;
746815
};

0 commit comments

Comments
 (0)