diff --git a/conformance_tests/tools/metrics/src/test_metric.cpp b/conformance_tests/tools/metrics/src/test_metric.cpp index 5e9887d2..537d251c 100644 --- a/conformance_tests/tools/metrics/src/test_metric.cpp +++ b/conformance_tests/tools/metrics/src/test_metric.cpp @@ -398,9 +398,11 @@ TEST_F( for (auto &test_metric_group : test_metric_groups) { lzt::activate_metric_groups(deviceh, 1, &test_metric_group); + uint32_t notifyEveryNReports = 100; + uint32_t samplingPeriod = 100000; zet_metric_streamer_handle_t metric_streamer_handle = lzt::metric_streamer_open_for_device(deviceh, test_metric_group, - nullptr, 100, 100000); + nullptr, notifyEveryNReports, samplingPeriod); auto report_size = lzt::metric_streamer_read_data_size(metric_streamer_handle, 1); lzt::metric_streamer_close(metric_streamer_handle); @@ -1231,6 +1233,7 @@ TEST_F( zetMetricStreamerTest, GivenValidMetricGroupWhenTimerBasedStreamerIsCreatedThenExpectStreamerToSucceed) { + notifyEveryNReports = 9000; for (auto device : devices) { ze_device_properties_t deviceProperties = { @@ -1321,6 +1324,197 @@ TEST_F( } } +TEST_F( + zetMetricStreamerTest, + GivenValidMetricGroupWhenTimerBasedStreamerIsCreatedThenExpectStreamerToNotifyEventAtProperTimeAndSucceed) { + + uint32_t notifyEveryNReports = 1000; + uint32_t samplingPeriod = 50000000; + for (auto device : devices) { + + ze_device_properties_t deviceProperties = { + ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES, nullptr}; + zeDeviceGetProperties(device, &deviceProperties); + + LOG_INFO << "test device name " << deviceProperties.name << " uuid " + << lzt::to_string(deviceProperties.uuid); + if (deviceProperties.flags & ZE_DEVICE_PROPERTY_FLAG_SUBDEVICE) { + LOG_INFO << "test subdevice id " << deviceProperties.subdeviceId; + } else { + LOG_INFO << "test device is a root device"; + } + + ze_command_queue_handle_t commandQueue = lzt::create_command_queue(device); + zet_command_list_handle_t commandList = lzt::create_command_list(device); + + auto metricGroupInfo = lzt::get_metric_group_info( + device, ZET_METRIC_GROUP_SAMPLING_TYPE_FLAG_TIME_BASED, true); + metricGroupInfo = lzt::optimize_metric_group_info_list(metricGroupInfo); + + void *a_buffer, *b_buffer, *c_buffer; + ze_group_count_t tg; + ze_kernel_handle_t function = get_matrix_multiplication_kernel( + device, &tg, &a_buffer, &b_buffer, &c_buffer, 8192); + zeCommandListAppendLaunchKernel(commandList, function, &tg, nullptr, 0, + nullptr); + lzt::close_command_list(commandList); + + for (auto groupInfo : metricGroupInfo) { + LOG_INFO << "test metricGroup name " << groupInfo.metricGroupName; + lzt::activate_metric_groups(device, 1, &groupInfo.metricGroupHandle); + + ze_event_handle_t eventHandle; + lzt::zeEventPool eventPool; + eventPool.create_event(eventHandle, ZE_EVENT_SCOPE_FLAG_HOST, + ZE_EVENT_SCOPE_FLAG_HOST); + + auto startTime = std::chrono::system_clock::now(); + zet_metric_streamer_handle_t metricStreamerHandle = + lzt::metric_streamer_open_for_device( + device, groupInfo.metricGroupHandle, eventHandle, + notifyEveryNReports, samplingPeriod); + ASSERT_NE(nullptr, metricStreamerHandle); + + lzt::execute_command_lists(commandQueue, 1, &commandList, nullptr); + lzt::synchronize(commandQueue, std::numeric_limits::max()); + + double minimumTimeBeforeEventIsExpected = notifyEveryNReports * (static_cast(samplingPeriod) / 1000000000); + // Initializing a 5% error buffer to prevent corner cases + double errorBuffer = 0.05 * minimumTimeBeforeEventIsExpected; + LOG_INFO << "minimumTimeBeforeEventIsExpected " << minimumTimeBeforeEventIsExpected; + LOG_INFO << "errorBuffer " << errorBuffer; + + ze_result_t eventResult; + while (true) { + auto currentTime = std::chrono::system_clock::now(); + std::chrono::duration elapsedSeconds = currentTime - startTime; + if (elapsedSeconds.count() >= (minimumTimeBeforeEventIsExpected - errorBuffer)) { + break; + } + eventResult = zeEventQueryStatus(eventHandle); + EXPECT_EQ(eventResult, ZE_RESULT_NOT_READY); + std::this_thread::sleep_for(std::chrono::milliseconds(1000)); + } + + // Sleep again for 2x the buffer time to ensure corner cases are avoided. 2x is the amount to maintain the timeline itself. + uint32_t sleep = std::ceil(2 * errorBuffer); + LOG_INFO << "additional sleep before expecting event to be ready " << sleep; + std::this_thread::sleep_for(std::chrono::seconds(sleep)); + + eventResult = zeEventQueryStatus(eventHandle); + EXPECT_EQ(eventResult, ZE_RESULT_SUCCESS); + + lzt::metric_streamer_close(metricStreamerHandle); + lzt::deactivate_metric_groups(device); + eventPool.destroy_event(eventHandle); + } + lzt::destroy_function(function); + lzt::free_memory(a_buffer); + lzt::free_memory(b_buffer); + lzt::free_memory(c_buffer); + lzt::reset_command_list(commandList); + lzt::destroy_command_queue(commandQueue); + lzt::destroy_command_list(commandList); + } +} + +TEST_F( + zetMetricStreamerTest, + GivenValidMetricGroupWhenTimerBasedStreamerIsCreatedThenExpectStreamerToNotifyEventAtProperTimeAndCorrectNumberOfReportsAreGenerated) { + + uint32_t notifyEveryNReports = 1000; + uint32_t samplingPeriod = 50000000; + for (auto device : devices) { + + ze_device_properties_t deviceProperties = { + ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES, nullptr}; + zeDeviceGetProperties(device, &deviceProperties); + + LOG_INFO << "test device name " << deviceProperties.name << " uuid " + << lzt::to_string(deviceProperties.uuid); + if (deviceProperties.flags & ZE_DEVICE_PROPERTY_FLAG_SUBDEVICE) { + LOG_INFO << "test subdevice id " << deviceProperties.subdeviceId; + } else { + LOG_INFO << "test device is a root device"; + } + + ze_command_queue_handle_t commandQueue = lzt::create_command_queue(device); + zet_command_list_handle_t commandList = lzt::create_command_list(device); + + auto metricGroupInfo = lzt::get_metric_group_info( + device, ZET_METRIC_GROUP_SAMPLING_TYPE_FLAG_TIME_BASED, true); + metricGroupInfo = lzt::optimize_metric_group_info_list(metricGroupInfo); + + void *a_buffer, *b_buffer, *c_buffer; + ze_group_count_t tg; + ze_kernel_handle_t function = get_matrix_multiplication_kernel( + device, &tg, &a_buffer, &b_buffer, &c_buffer, 8192); + zeCommandListAppendLaunchKernel(commandList, function, &tg, nullptr, 0, + nullptr); + lzt::close_command_list(commandList); + + for (auto groupInfo : metricGroupInfo) { + LOG_INFO << "test metricGroup name " << groupInfo.metricGroupName; + lzt::activate_metric_groups(device, 1, &groupInfo.metricGroupHandle); + + ze_event_handle_t eventHandle; + lzt::zeEventPool eventPool; + eventPool.create_event(eventHandle, ZE_EVENT_SCOPE_FLAG_HOST, + ZE_EVENT_SCOPE_FLAG_HOST); + + auto startTime = std::chrono::system_clock::now(); + zet_metric_streamer_handle_t metricStreamerHandle = + lzt::metric_streamer_open_for_device( + device, groupInfo.metricGroupHandle, eventHandle, + notifyEveryNReports, samplingPeriod); + ASSERT_NE(nullptr, metricStreamerHandle); + + lzt::execute_command_lists(commandQueue, 1, &commandList, nullptr); + lzt::synchronize(commandQueue, std::numeric_limits::max()); + + double minimumTimeBeforeEventIsExpected = notifyEveryNReports * (static_cast(samplingPeriod) / 1000000000); + // Initializing a 5% error buffer to prevent corner cases + double errorBuffer = 0.05 * minimumTimeBeforeEventIsExpected; + LOG_INFO << "minimumTimeBeforeEventIsExpected " << minimumTimeBeforeEventIsExpected; + LOG_INFO << "errorBuffer " << errorBuffer; + + // Sleep untill event is generated. + auto currentTime = std::chrono::system_clock::now(); + std::chrono::duration elapsedSeconds = currentTime - startTime; + uint32_t timeLeft = std::ceil(minimumTimeBeforeEventIsExpected + errorBuffer - elapsedSeconds.count()); + if (timeLeft > 0) { + LOG_INFO << "additional sleep before expecting event to be ready " << timeLeft; + std::this_thread::sleep_for(std::chrono::seconds(timeLeft)); + } + + ze_result_t eventResult; + eventResult = zeEventQueryStatus(eventHandle); + EXPECT_EQ(eventResult, ZE_RESULT_SUCCESS); + + size_t oneReportSize, allReportsSize; + oneReportSize = + lzt::metric_streamer_read_data_size(metricStreamerHandle, 1); + allReportsSize = lzt::metric_streamer_read_data_size( + metricStreamerHandle, UINT32_MAX); + LOG_DEBUG << "Event triggered. Single report size: " << oneReportSize + << ". All reports size:" << allReportsSize; + + EXPECT_GE(allReportsSize / oneReportSize, notifyEveryNReports); + + lzt::metric_streamer_close(metricStreamerHandle); + lzt::deactivate_metric_groups(device); + eventPool.destroy_event(eventHandle); + } + lzt::destroy_function(function); + lzt::free_memory(a_buffer); + lzt::free_memory(b_buffer); + lzt::free_memory(c_buffer); + lzt::reset_command_list(commandList); + lzt::destroy_command_queue(commandQueue); + lzt::destroy_command_list(commandList); + } +} + void metric_validate_stall_sampling_data( std::vector &metricProperties, std::vector &totalMetricValues, @@ -2330,8 +2524,8 @@ void run_multi_device_streamer_test( eventPool.create_event(eventHandle_1, ZE_EVENT_SCOPE_FLAG_HOST, ZE_EVENT_SCOPE_FLAG_HOST); - auto notifyEveryNReports = 3000; - auto samplingPeriod = 1000000; + uint32_t notifyEveryNReports = 3000; + uint32_t samplingPeriod = 1000000; auto metricStreamerHandle_0 = lzt::metric_streamer_open_for_device( device_0, groupInfo_0.metricGroupHandle, eventHandle_0, notifyEveryNReports, samplingPeriod); diff --git a/utils/test_harness/tools/include/test_harness_metric.hpp b/utils/test_harness/tools/include/test_harness_metric.hpp index 45d2d41c..e5796de5 100644 --- a/utils/test_harness/tools/include/test_harness_metric.hpp +++ b/utils/test_harness/tools/include/test_harness_metric.hpp @@ -86,8 +86,8 @@ void metric_query_get_data(zet_metric_query_handle_t metricQueryHandle, std::vector *metricData); zet_metric_streamer_handle_t metric_streamer_open_for_device( ze_device_handle_t device, zet_metric_group_handle_t matchedGroupHandle, - ze_event_handle_t eventHandle, uint32_t notifyEveryNReports, - uint32_t samplingPeriod); + ze_event_handle_t eventHandle, uint32_t ¬ifyEveryNReports, + uint32_t &samplingPeriod); zet_metric_streamer_handle_t metric_streamer_open(zet_metric_group_handle_t matchedGroupHandle, ze_event_handle_t eventHandle, diff --git a/utils/test_harness/tools/src/test_harness_metric.cpp b/utils/test_harness/tools/src/test_harness_metric.cpp index 59a8e308..c0b15f0b 100644 --- a/utils/test_harness/tools/src/test_harness_metric.cpp +++ b/utils/test_harness/tools/src/test_harness_metric.cpp @@ -516,8 +516,8 @@ metric_streamer_open(zet_metric_group_handle_t matchedGroupHandle, zet_metric_streamer_handle_t metric_streamer_open_for_device( ze_device_handle_t device, zet_metric_group_handle_t matchedGroupHandle, - ze_event_handle_t eventHandle, uint32_t notifyEveryNReports, - uint32_t samplingPeriod) { + ze_event_handle_t eventHandle, uint32_t ¬ifyEveryNReports, + uint32_t &samplingPeriod) { zet_metric_streamer_handle_t metricStreamerHandle = nullptr; zet_metric_streamer_desc_t metricStreamerDesc = { ZET_STRUCTURE_TYPE_METRIC_STREAMER_DESC, nullptr, notifyEveryNReports, @@ -527,6 +527,8 @@ zet_metric_streamer_handle_t metric_streamer_open_for_device( matchedGroupHandle, &metricStreamerDesc, eventHandle, &metricStreamerHandle)); EXPECT_NE(nullptr, metricStreamerHandle); + notifyEveryNReports = metricStreamerDesc.notifyEveryNReports; + samplingPeriod = metricStreamerDesc.samplingPeriod; return metricStreamerHandle; }