intel/compute-runtime
1
0
Fork 0
mirror of https://github.com/intel/compute-runtime.git synced 2025-12-21 09:14:47 +08:00
Code Issues Packages Projects Releases Wiki Activity

test: add counter based checker and event pool option

Browse Source 
Signed-off-by: Zbigniew Zdanowicz <zbigniew.zdanowicz@intel.com>
This commit is contained in:
Zbigniew Zdanowicz
2024-03-05 00:45:31 +00:00
committed by Compute-Runtime-Automation
parent 7bb23818e6
commit 178dfe0860
9 changed files with 78 additions and 32 deletions
Download Patch File Download Diff File Expand all files Collapse all files

57
level_zero/core/test/black_box_tests/common/zello_common.cpp
View File

@@ -19,6 +19,13 @@
namespace LevelZeroBlackBoxTests {
struct LoadedDriverExtensions {
std::vector<ze_driver_extension_properties_t> extensions;
bool loaded = false;
};
static LoadedDriverExtensions driverExtensions;
bool verbose;
bool isParamEnabled(int argc, char *argv[], const char *shortName, const char *longName) {
@@ -354,16 +361,23 @@ void createEventPoolAndEvents(ze_context_handle_t &context,
ze_device_handle_t &device,
ze_event_pool_handle_t &eventPool,
ze_event_pool_flags_t poolFlag,
bool counterEvents,
ze_event_pool_counter_based_exp_flags_t poolCounterFlag,
uint32_t poolSize,
ze_event_handle_t *events,
ze_event_scope_flags_t signalScope,
ze_event_scope_flags_t waitScope) {
ze_event_pool_counter_based_exp_desc_t counterPoolDesc{ZE_STRUCTURE_TYPE_COUNTER_BASED_EVENT_POOL_EXP_DESC};
counterPoolDesc.flags = poolCounterFlag;
ze_event_pool_desc_t eventPoolDesc{ZE_STRUCTURE_TYPE_EVENT_POOL_DESC};
ze_event_desc_t eventDesc = {ZE_STRUCTURE_TYPE_EVENT_DESC};
eventPoolDesc.count = poolSize;
eventPoolDesc.flags = poolFlag;
if (counterEvents) {
eventPoolDesc.pNext = &counterPoolDesc;
}
SUCCESS_OR_TERMINATE(zeEventPoolCreate(context, &eventPoolDesc, 1, &device, &eventPool));
ze_event_desc_t eventDesc = {ZE_STRUCTURE_TYPE_EVENT_DESC};
for (uint32_t i = 0; i < poolSize; i++) {
eventDesc.index = i;
eventDesc.signal = signalScope;
@@ -372,6 +386,19 @@ void createEventPoolAndEvents(ze_context_handle_t &context,
}
}
bool counterBasedEventsExtensionPresent(ze_driver_handle_t &driverHandle) {
std::string cbEventsExtensionString("ZE_experimental_event_pool_counter_based");
ze_driver_extension_properties_t cbEventsExtension{};
strncpy(cbEventsExtension.name, cbEventsExtensionString.c_str(), cbEventsExtensionString.size());
cbEventsExtension.version = ZE_EVENT_POOL_COUNTER_BASED_EXP_VERSION_CURRENT;
std::vector<ze_driver_extension_properties_t> extensionsToCheck;
extensionsToCheck.push_back(cbEventsExtension);
return LevelZeroBlackBoxTests::checkExtensionIsPresent(driverHandle, extensionsToCheck);
}
std::vector<ze_device_handle_t> zelloGetSubDevices(ze_device_handle_t &device, uint32_t &subDevCount) {
uint32_t deviceCount = 0;
std::vector<ze_device_handle_t> subdevs(deviceCount, nullptr);
@@ -673,28 +700,40 @@ void printBuildLog(const char *strLog) {
<< strLog << std::endl;
}
bool checkExtensionIsPresent(ze_driver_handle_t &driverHandle, std::vector<ze_driver_extension_properties_t> &extensionsToCheck) {
uint32_t numMatchedExtensions = 0;
void loadDriverExtensions(ze_driver_handle_t &driverHandle, std::vector<ze_driver_extension_properties_t> &driverExtensions) {
uint32_t extensionsCount = 0;
SUCCESS_OR_TERMINATE(zeDriverGetExtensionProperties(driverHandle, &extensionsCount, nullptr));
if (extensionsCount == 0) {
std::cerr << "No extensions supported on this driver" << std::endl;
return false;
driverExtensions.resize(extensionsCount);
return;
}
std::vector<ze_driver_extension_properties_t> extensionsSupported(extensionsCount);
SUCCESS_OR_TERMINATE(zeDriverGetExtensionProperties(driverHandle, &extensionsCount, extensionsSupported.data()));
driverExtensions.resize(extensionsCount);
SUCCESS_OR_TERMINATE(zeDriverGetExtensionProperties(driverHandle, &extensionsCount, driverExtensions.data()));
for (uint32_t i = 0; i < extensionsCount; i++) {
uint32_t supportedVersion = extensionsSupported[i].version;
uint32_t supportedVersion = driverExtensions[i].version;
if (verbose) {
std::cout << "Extension #" << i << " name: " << extensionsSupported[i].name << " version: " << ZE_MAJOR_VERSION(supportedVersion) << "." << ZE_MINOR_VERSION(supportedVersion) << std::endl;
std::cout << "Extension #" << i << " name: " << driverExtensions[i].name << " version: " << ZE_MAJOR_VERSION(supportedVersion) << "." << ZE_MINOR_VERSION(supportedVersion) << std::endl;
}
}
}
bool checkExtensionIsPresent(ze_driver_handle_t &driverHandle, std::vector<ze_driver_extension_properties_t> &extensionsToCheck) {
uint32_t numMatchedExtensions = 0;
if (!driverExtensions.loaded) {
loadDriverExtensions(driverHandle, driverExtensions.extensions);
driverExtensions.loaded = true;
}
for (uint32_t i = 0; i < driverExtensions.extensions.size(); i++) {
for (const auto &checkedExtension : extensionsToCheck) {
std::string checkedExtensionName = checkedExtension.name;
if (strncmp(extensionsSupported[i].name, checkedExtensionName.c_str(), checkedExtensionName.size()) == 0) {
if (strncmp(driverExtensions.extensions[i].name, checkedExtensionName.c_str(), checkedExtensionName.size()) == 0) {
uint32_t version = checkedExtension.version;
uint32_t supportedVersion = driverExtensions.extensions[i].version;
if (verbose) {
std::cout << "Checked extension: " << checkedExtensionName << " found in the driver." << std::endl;
}

6
level_zero/core/test/black_box_tests/common/zello_common.h
View File

@@ -106,11 +106,15 @@ void createEventPoolAndEvents(ze_context_handle_t &context,
ze_device_handle_t &device,
ze_event_pool_handle_t &eventPool,
ze_event_pool_flags_t poolFlag,
bool counterEvents,
ze_event_pool_counter_based_exp_flags_t poolCounterFlag,
uint32_t poolSize,
ze_event_handle_t *events,
ze_event_scope_flags_t signalScope,
ze_event_scope_flags_t waitScope);
bool counterBasedEventsExtensionPresent(ze_driver_handle_t &driverHandle);
std::vector<ze_device_handle_t> zelloGetSubDevices(ze_device_handle_t &device, uint32_t &subDevCount);
std::vector<ze_device_handle_t> zelloInitContextAndGetDevices(ze_context_handle_t &context, ze_driver_handle_t &driverHandle);
@@ -277,6 +281,8 @@ void printBuildLog(std::string &buildLog);
void printBuildLog(const char *strLog);
void loadDriverExtensions(ze_driver_handle_t &driverHandle, std::vector<ze_driver_extension_properties_t> &driverExtensions);
bool checkExtensionIsPresent(ze_driver_handle_t &driverHandle, std::vector<ze_driver_extension_properties_t> &extensionsToCheck);
} // namespace LevelZeroBlackBoxTests

2
level_zero/core/test/black_box_tests/zello_atomic_inc.cpp
View File

@@ -97,7 +97,7 @@ void executeKernelAndValidate(ze_context_handle_t &context, ze_device_handle_t &
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> events(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP,
ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, false, 0,
numEvents, events.data(),
ZE_EVENT_SCOPE_FLAG_DEVICE,
0);

16
level_zero/core/test/black_box_tests/zello_commandlist_immediate.cpp
View File

@@ -54,12 +54,12 @@ void testAppendMemoryCopy(ze_context_handle_t &context, ze_device_handle_t &devi
if (!useSyncCmdQ) {
if (sharedEvent == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event = sharedEvent;
}
if (sharedEvent2 == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event2 = sharedEvent2;
}
@@ -172,12 +172,12 @@ void testAppendMemoryCopyRegion(ze_context_handle_t &context, ze_device_handle_t
if (!useSyncCmdQ) {
// Create Event Pool and kernel launch event
if (sharedEvent == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event = sharedEvent;
}
if (sharedEvent2 == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event2 = sharedEvent2;
}
@@ -342,12 +342,12 @@ void testAppendGpuKernel(ze_context_handle_t &context, ze_device_handle_t &devic
if (!useSyncCmdQ) {
// Create Event Pool and kernel launch event
if (sharedEvent == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event = sharedEvent;
}
if (sharedEvent2 == nullptr) {
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
event2 = sharedEvent2;
}
@@ -453,8 +453,8 @@ int main(int argc, char *argv[]) {
if (!useSyncQueue && eventPoolShared) {
// Create Event Pool and kernel launch event
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device0, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device0, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device0, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device0, eventPool2, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event2, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
}
ze_command_list_handle_t cmdList = nullptr;

9
level_zero/core/test/black_box_tests/zello_events.cpp
View File

@@ -61,12 +61,12 @@ bool testEventsDeviceSignalDeviceWait(ze_context_handle_t &context, ze_device_ha
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> deviceEvents(numEvents), hostEvents(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPoolDevice,
0,
0, false, 0,
numEvents, deviceEvents.data(),
ZE_EVENT_SCOPE_FLAG_SUBDEVICE,
0);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPoolHost,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0,
numEvents, hostEvents.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);
@@ -152,7 +152,7 @@ bool testEventsDeviceSignalHostWait(ze_context_handle_t &context, ze_device_hand
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> events(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0,
numEvents, events.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);
@@ -211,7 +211,7 @@ bool testEventsDeviceSignalHostWaitWithNonZeroOrdinal(ze_context_handle_t &conte
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> events(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0,
numEvents, events.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);
@@ -272,6 +272,7 @@ bool testEventsHostSignalHostWait(ze_context_handle_t &context, ze_device_handle
std::vector<ze_event_handle_t> events(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE | ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP,
false, 0,
numEvents, events.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);

6
level_zero/core/test/black_box_tests/zello_immediate.cpp
View File

@@ -65,12 +65,12 @@ void testCopyBetweenHostMemAndDeviceMem(ze_context_handle_t &context, ze_device_
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> deviceEvents(numEvents), hostEvents(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPoolDevice,
0,
0, false, 0,
numEvents, deviceEvents.data(),
ZE_EVENT_SCOPE_FLAG_SUBDEVICE,
0);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPoolHost,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0,
numEvents, hostEvents.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);
@@ -150,7 +150,7 @@ void executeGpuKernelAndValidate(ze_context_handle_t &context, ze_device_handle_
uint32_t numEvents = 2;
std::vector<ze_event_handle_t> hostEvents(numEvents);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPoolHost,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE,
ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0,
numEvents, hostEvents.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
0);

4
level_zero/core/test/black_box_tests/zello_sandbox.cpp
View File

@@ -115,7 +115,7 @@ void executeImmediateAndRegularCommandLists(ze_context_handle_t &context, ze_dev
ze_event_pool_handle_t eventPool;
ze_event_handle_t events[3];
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 3, events, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 3, events, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
SUCCESS_OR_TERMINATE(zeCommandListAppendMemoryCopy(cmdList, deviceMemory, sourceSystemMemory.data(), regularCmdlistBufSize, events[1], 0, nullptr));
SUCCESS_OR_TERMINATE(zeCommandListAppendMemoryCopy(cmdList, destSystemMemory.data(), deviceMemory, regularCmdlistBufSize, events[2], 1, &events[1]));
@@ -208,7 +208,7 @@ void executeMemoryTransferAndValidate(ze_context_handle_t &context, ze_device_ha
std::vector<ze_event_handle_t> events(numEvents);
ze_event_pool_flags_t eventPoolFlags = flags;
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
eventPoolFlags,
eventPoolFlags, false, 0,
numEvents, events.data(),
ZE_EVENT_SCOPE_FLAG_HOST,
ZE_EVENT_SCOPE_FLAG_HOST);

2
level_zero/core/test/black_box_tests/zello_scratch.cpp
View File

@@ -50,7 +50,7 @@ void executeGpuKernelAndValidate(ze_context_handle_t &context,
cmdQueueDesc.index = 0;
if (useAsync) {
cmdQueueDesc.mode = ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS;
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_HOST_VISIBLE, false, 0, 1, &event, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
} else {
cmdQueueDesc.mode = ZE_COMMAND_QUEUE_MODE_SYNCHRONOUS;
}

8
level_zero/core/test/black_box_tests/zello_timestamp.cpp
View File

@@ -228,7 +228,7 @@ bool testKernelTimestampHostQuery(int argc, char *argv[],
ze_event_pool_handle_t eventPool;
ze_event_handle_t kernelTsEvent;
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, false, 0, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
SUCCESS_OR_TERMINATE(zeCommandListAppendLaunchKernel(cmdList, kernel, &dispatchTraits, kernelTsEvent, 0, nullptr));
SUCCESS_OR_TERMINATE(zeCommandListClose(cmdList));
@@ -336,7 +336,7 @@ bool testKernelTimestampAppendQuery(ze_context_handle_t &context,
ze_event_pool_handle_t eventPool;
ze_event_handle_t kernelTsEvent;
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, false, 0, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
SUCCESS_OR_TERMINATE(zeCommandListAppendLaunchKernel(cmdList, kernel, &dispatchTraits, kernelTsEvent, 0, nullptr));
SUCCESS_OR_TERMINATE(zeCommandListAppendBarrier(cmdList, nullptr, 0u, nullptr));
@@ -485,7 +485,7 @@ bool testKernelTimestampMapToHostTimescale(int argc, char *argv[],
dispatchTraits.groupCountY = 1u;
dispatchTraits.groupCountZ = 1u;
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool, ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP, false, 0, 1, &kernelTsEvent, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST);
SUCCESS_OR_TERMINATE(zeCommandListAppendLaunchKernel(cmdList, kernel, &dispatchTraits, kernelTsEvent, 0, nullptr));
return true;
@@ -620,7 +620,7 @@ bool testKernelMappedTimestampMap(int argc, char *argv[],
ze_event_handle_t kernelTsEvent[maxEventUsageCount];
LevelZeroBlackBoxTests::createEventPoolAndEvents(context, device, eventPool,
(ZE_EVENT_POOL_FLAG_HOST_VISIBLE | ZE_EVENT_POOL_FLAG_KERNEL_MAPPED_TIMESTAMP),
(ZE_EVENT_POOL_FLAG_HOST_VISIBLE | ZE_EVENT_POOL_FLAG_KERNEL_MAPPED_TIMESTAMP), false, 0,
maxEventUsageCount, kernelTsEvent,
ZE_EVENT_SCOPE_FLAG_DEVICE, ZE_EVENT_SCOPE_FLAG_HOST);