From 321f285fd9db573c481937383ee6039d8c211a3f Mon Sep 17 00:00:00 2001 From: Compute-Runtime-Validation Date: Tue, 22 Jul 2025 21:04:55 +0200 Subject: [PATCH] Revert "fix: correct global device synchronization" This reverts commit 8bdc479fe77d66de60a07037ccef13068a504f7d. Signed-off-by: Compute-Runtime-Validation --- level_zero/core/source/device/device_imp.cpp | 32 ++-- .../core/test/black_box_tests/CMakeLists.txt | 1 - .../black_box_tests/common/zello_common.cpp | 14 +- .../black_box_tests/common/zello_common.h | 10 +- .../black_box_tests/common/zello_compile.cpp | 27 --- .../black_box_tests/common/zello_compile.h | 1 - .../l0_blackbox_runner_configs.yml | 5 - .../test/black_box_tests/zello_arg_slm.cpp | 155 ------------------ .../sources/device/test_l0_device.cpp | 47 ------ 9 files changed, 18 insertions(+), 274 deletions(-) delete mode 100644 level_zero/core/test/black_box_tests/zello_arg_slm.cpp diff --git a/level_zero/core/source/device/device_imp.cpp b/level_zero/core/source/device/device_imp.cpp index 32d4429242..e72a18104e 100644 --- a/level_zero/core/source/device/device_imp.cpp +++ b/level_zero/core/source/device/device_imp.cpp @@ -2247,27 +2247,23 @@ uint32_t DeviceImp::getEventMaxKernelCount() const { ze_result_t DeviceImp::synchronize() { for (auto &engine : neoDevice->getAllEngines()) { - if (engine.commandStreamReceiver->isInitialized()) { - auto waitStatus = engine.commandStreamReceiver->waitForTaskCountWithKmdNotifyFallback( - engine.commandStreamReceiver->peekTaskCount(), - engine.commandStreamReceiver->obtainCurrentFlushStamp(), - false, - NEO::QueueThrottle::MEDIUM); - if (waitStatus == NEO::WaitStatus::gpuHang) { - return ZE_RESULT_ERROR_DEVICE_LOST; - } + auto waitStatus = engine.commandStreamReceiver->waitForTaskCountWithKmdNotifyFallback( + engine.commandStreamReceiver->peekTaskCount(), + engine.commandStreamReceiver->obtainCurrentFlushStamp(), + false, + NEO::QueueThrottle::MEDIUM); + if (waitStatus == NEO::WaitStatus::gpuHang) { + return ZE_RESULT_ERROR_DEVICE_LOST; } } for (auto &secondaryCsr : neoDevice->getSecondaryCsrs()) { - if (secondaryCsr->isInitialized()) { - auto waitStatus = secondaryCsr->waitForTaskCountWithKmdNotifyFallback( - secondaryCsr->peekTaskCount(), - secondaryCsr->obtainCurrentFlushStamp(), - false, - NEO::QueueThrottle::MEDIUM); - if (waitStatus == NEO::WaitStatus::gpuHang) { - return ZE_RESULT_ERROR_DEVICE_LOST; - } + auto waitStatus = secondaryCsr->waitForTaskCountWithKmdNotifyFallback( + secondaryCsr->peekTaskCount(), + secondaryCsr->obtainCurrentFlushStamp(), + false, + NEO::QueueThrottle::MEDIUM); + if (waitStatus == NEO::WaitStatus::gpuHang) { + return ZE_RESULT_ERROR_DEVICE_LOST; } } diff --git a/level_zero/core/test/black_box_tests/CMakeLists.txt b/level_zero/core/test/black_box_tests/CMakeLists.txt index dfef8e3500..d99092fd5d 100644 --- a/level_zero/core/test/black_box_tests/CMakeLists.txt +++ b/level_zero/core/test/black_box_tests/CMakeLists.txt @@ -38,7 +38,6 @@ target_include_directories(${L0_BLACK_BOX_TEST_SHARED_LIB} PUBLIC set_target_properties(${L0_BLACK_BOX_TEST_SHARED_LIB} PROPERTIES FOLDER ${L0_BLACK_BOX_TEST_PROJECT_FOLDER}) set(TEST_TARGETS - zello_arg_slm zello_atomic_inc zello_bindless_kernel zello_commandlist_immediate diff --git a/level_zero/core/test/black_box_tests/common/zello_common.cpp b/level_zero/core/test/black_box_tests/common/zello_common.cpp index f4488e5f1a..c1f4e2f833 100644 --- a/level_zero/core/test/black_box_tests/common/zello_common.cpp +++ b/level_zero/core/test/black_box_tests/common/zello_common.cpp @@ -20,11 +20,6 @@ namespace LevelZeroBlackBoxTests { decltype(&zerDriverGetDefaultContext) zerDriverGetDefaultContextFunc = nullptr; -decltype(&zeDeviceSynchronize) zeDeviceSynchronizeFunc = nullptr; -decltype(&zeCommandListAppendLaunchKernelWithArguments) zeCommandListAppendLaunchKernelWithArgumentsFunc = nullptr; -decltype(&zerIdentifierTranslateToDeviceHandle) zerIdentifierTranslateToDeviceHandleFunc = nullptr; -decltype(&zerDeviceTranslateToIdentifier) zerDeviceTranslateToIdentifierFunc = nullptr; -decltype(&zerDriverGetLastErrorDescription) zerDriverGetLastErrorDescriptionFunc = nullptr; struct LoadedDriverExtensions { std::vector extensions; @@ -266,7 +261,7 @@ void getErrorMax(int argc, char *argv[]) { overrideErrorMax = getParamValue(argc, argv, "-em", "--errorMax", 0); } -void printResult(bool aubMode, bool outputValidationSuccessful, const std::string_view blackBoxName, const std::string_view currentTest) { +void printResult(bool aubMode, bool outputValidationSuccessful, const std::string &blackBoxName, const std::string ¤tTest) { std::cout << std::endl << blackBoxName; if (!currentTest.empty()) { @@ -285,7 +280,7 @@ void printResult(bool aubMode, bool outputValidationSuccessful, const std::strin } } -void printResult(bool aubMode, bool outputValidationSuccessful, const std::string_view blackBoxName) { +void printResult(bool aubMode, bool outputValidationSuccessful, const std::string &blackBoxName) { std::string currentTest{}; printResult(aubMode, outputValidationSuccessful, blackBoxName, currentTest); } @@ -459,11 +454,6 @@ std::vector zelloInitContextAndGetDevices(ze_context_handle_ SUCCESS_OR_TERMINATE(zeDriverGet(&driverCount, &driverHandle)); SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zerDriverGetDefaultContext", reinterpret_cast(&zerDriverGetDefaultContextFunc))); - SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zeDeviceSynchronize", reinterpret_cast(&zeDeviceSynchronizeFunc))); - SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zeCommandListAppendLaunchKernelWithArguments", reinterpret_cast(&zeCommandListAppendLaunchKernelWithArgumentsFunc))); - SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zerIdentifierTranslateToDeviceHandle", reinterpret_cast(&zerIdentifierTranslateToDeviceHandleFunc))); - SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zerDeviceTranslateToIdentifier", reinterpret_cast(&zerDeviceTranslateToIdentifierFunc))); - SUCCESS_OR_TERMINATE(zeDriverGetExtensionFunctionAddress(driverHandle, "zerDriverGetLastErrorDescription", reinterpret_cast(&zerDriverGetLastErrorDescriptionFunc))); context = zerDriverGetDefaultContextFunc(); if (!context) { diff --git a/level_zero/core/test/black_box_tests/common/zello_common.h b/level_zero/core/test/black_box_tests/common/zello_common.h index 2f3cdcb061..0f3eb9b4d0 100644 --- a/level_zero/core/test/black_box_tests/common/zello_common.h +++ b/level_zero/core/test/black_box_tests/common/zello_common.h @@ -23,12 +23,6 @@ namespace LevelZeroBlackBoxTests { template inline void validate(ResulT result, const char *message); extern decltype(&zerDriverGetDefaultContext) zerDriverGetDefaultContextFunc; -extern decltype(&zeDeviceSynchronize) zeDeviceSynchronizeFunc; -extern decltype(&zeCommandListAppendLaunchKernelWithArguments) zeCommandListAppendLaunchKernelWithArgumentsFunc; -extern decltype(&zerIdentifierTranslateToDeviceHandle) zerIdentifierTranslateToDeviceHandleFunc; -extern decltype(&zerDeviceTranslateToIdentifier) zerDeviceTranslateToIdentifierFunc; -extern decltype(&zerDriverGetLastErrorDescription) zerDriverGetLastErrorDescriptionFunc; - } // namespace LevelZeroBlackBoxTests #define SUCCESS_OR_TERMINATE(CALL) LevelZeroBlackBoxTests::validate(CALL, #CALL) @@ -94,9 +88,9 @@ uint32_t getBufferLength(int argc, char *argv[], uint32_t defaultLength); void getErrorMax(int argc, char *argv[]); -void printResult(bool aubMode, bool outputValidationSuccessful, const std::string_view blackBoxName, const std::string_view currentTest); +void printResult(bool aubMode, bool outputValidationSuccessful, const std::string &blackBoxName, const std::string ¤tTest); -void printResult(bool aubMode, bool outputValidationSuccessful, const std::string_view blackBoxName); +void printResult(bool aubMode, bool outputValidationSuccessful, const std::string &blackBoxName); uint32_t getCommandQueueOrdinal(ze_device_handle_t &device, bool useCooperativeFlag); diff --git a/level_zero/core/test/black_box_tests/common/zello_compile.cpp b/level_zero/core/test/black_box_tests/common/zello_compile.cpp index 98754dec96..7a0f01a154 100644 --- a/level_zero/core/test/black_box_tests/common/zello_compile.cpp +++ b/level_zero/core/test/black_box_tests/common/zello_compile.cpp @@ -146,33 +146,6 @@ std::vector compileToNative(const std::string &src, const std::string & oclocFreeOutput(&numOutputs, &outputs, &ouputLengths, &outputNames); return ret; } -const char *slmArgKernelSrc = R"===( -__kernel void test_arg_slm( - __global unsigned int *outputSums, // Output array for sums (global memory) - __local unsigned int *localIdArray, // Local array for local IDs (shared memory) - __local unsigned int *globalIdArray // Local array for global IDs (shared memory) -) { - // Each work-item stores its local and global ID in local memory - localIdArray[get_local_id(0)] = get_local_id(0); - globalIdArray[get_local_id(0)] = get_global_id(0); - - // Synchronize all work-items in the group - barrier(CLK_LOCAL_MEM_FENCE); - - // Only the first work-item in the group performs the reduction - if(get_local_id(0) == 0){ - unsigned int sumLocalIds = 0; - unsigned int sumGlobalIds = 0; - for(int i = 0; i < get_local_size(0); ++i){ - sumLocalIds += localIdArray[i]; - sumGlobalIds += globalIdArray[i]; - } - // Store the results in the output array - outputSums[get_group_id(0)*2] = sumLocalIds; - outputSums[get_group_id(0)*2+1] = sumGlobalIds; - } -} -)==="; const char *memcpyBytesTestKernelSrc = R"===( kernel void memcpy_bytes(__global char *dst, const __global char *src) { diff --git a/level_zero/core/test/black_box_tests/common/zello_compile.h b/level_zero/core/test/black_box_tests/common/zello_compile.h index 59f2d4d29a..d22be6a44a 100644 --- a/level_zero/core/test/black_box_tests/common/zello_compile.h +++ b/level_zero/core/test/black_box_tests/common/zello_compile.h @@ -19,7 +19,6 @@ std::vector compileToSpirV(const std::string &src, const std::string &o std::vector compileToSpirV(const std::string &src, const std::string &options, const std::string &device, std::string &outCompilerLog); std::vector compileToNative(const std::string &src, const std::string &deviceName, const std::string &revisionId, const std::string &options, const std::string &internalOptions, const std::string &statefulMode, std::string &outCompilerLog); -extern const char *slmArgKernelSrc; extern const char *memcpyBytesTestKernelSrc; extern const char *memcpyBytesWithPrintfTestKernelSrc; diff --git a/level_zero/core/test/black_box_tests/l0_blackbox_runner_configs.yml b/level_zero/core/test/black_box_tests/l0_blackbox_runner_configs.yml index 0448a5e38c..8100d6259d 100644 --- a/level_zero/core/test/black_box_tests/l0_blackbox_runner_configs.yml +++ b/level_zero/core/test/black_box_tests/l0_blackbox_runner_configs.yml @@ -46,11 +46,6 @@ _default_config: params: - --verbose -zello_arg_slm: - bmg: - dg2: - pvc.b0: - zello_atomic_inc: dg2: pvc.b0: diff --git a/level_zero/core/test/black_box_tests/zello_arg_slm.cpp b/level_zero/core/test/black_box_tests/zello_arg_slm.cpp deleted file mode 100644 index 653adec8b8..0000000000 --- a/level_zero/core/test/black_box_tests/zello_arg_slm.cpp +++ /dev/null @@ -1,155 +0,0 @@ -/* - * Copyright (C) 2025 Intel Corporation - * - * SPDX-License-Identifier: MIT - * - */ - -#include "zello_common.h" -#include "zello_compile.h" - -#include - -constexpr std::string_view blackBoxName = "Zello Arg Slm"; - -void executeKernelAndValidate(ze_context_handle_t context, uint32_t deviceIdentfier, - bool &outputValidationSuccessful) { - - if (LevelZeroBlackBoxTests::verbose) { - std::cout << "Testing for device " << deviceIdentfier << std::endl; - } - ze_command_list_handle_t cmdList; - auto device = LevelZeroBlackBoxTests::zerIdentifierTranslateToDeviceHandleFunc(deviceIdentfier); - SUCCESS_OR_TERMINATE(zeCommandListCreateImmediate(context, device, &defaultCommandQueueDesc, &cmdList)); - - constexpr ze_group_count_t groupCounts{16, 1, 1}; - - // Create output buffer - void *dstBuffer = nullptr; - constexpr size_t allocSize = groupCounts.groupCountX * sizeof(uint32_t) * 2; - SUCCESS_OR_TERMINATE(zeMemAllocShared(context, &defaultDeviceMemDesc, &defaultHostMemDesc, allocSize, sizeof(uint32_t), device, &dstBuffer)); - - std::string buildLog; - auto spirV = LevelZeroBlackBoxTests::compileToSpirV(LevelZeroBlackBoxTests::slmArgKernelSrc, "", buildLog); - LevelZeroBlackBoxTests::printBuildLog(buildLog); - SUCCESS_OR_TERMINATE((0 == spirV.size())); - - ze_module_handle_t module = nullptr; - ze_kernel_handle_t kernel = nullptr; - - ze_module_desc_t moduleDesc = {ZE_STRUCTURE_TYPE_MODULE_DESC}; - ze_module_build_log_handle_t buildlog; - moduleDesc.format = ZE_MODULE_FORMAT_IL_SPIRV; - moduleDesc.pInputModule = spirV.data(); - moduleDesc.inputSize = spirV.size(); - moduleDesc.pBuildFlags = ""; - - if (zeModuleCreate(context, device, &moduleDesc, &module, &buildlog) != ZE_RESULT_SUCCESS) { - size_t szLog = 0; - zeModuleBuildLogGetString(buildlog, &szLog, nullptr); - - std::vector strLog(szLog + 1, 0); - zeModuleBuildLogGetString(buildlog, &szLog, strLog.data()); - LevelZeroBlackBoxTests::printBuildLog(strLog.data()); - - SUCCESS_OR_TERMINATE(zeModuleBuildLogDestroy(buildlog)); - std::cerr << std::endl - << blackBoxName << " Results validation FAILED. Module creation error." - << std::endl; - SUCCESS_OR_TERMINATE_BOOL(false); - } - SUCCESS_OR_TERMINATE(zeModuleBuildLogDestroy(buildlog)); - - ze_kernel_desc_t kernelDesc = {ZE_STRUCTURE_TYPE_KERNEL_DESC}; - kernelDesc.pKernelName = "test_arg_slm"; - SUCCESS_OR_TERMINATE(zeKernelCreate(module, &kernelDesc, &kernel)); - ze_kernel_properties_t kernProps = {ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES}; - SUCCESS_OR_TERMINATE(zeKernelGetProperties(kernel, &kernProps)); - LevelZeroBlackBoxTests::printKernelProperties(kernProps, kernelDesc.pKernelName); - - for (auto groupSize : {64u, 128u, 256u}) { - - // Initialize memory - constexpr uint8_t initValue = 77; - zeCommandListAppendMemoryFill(cmdList, dstBuffer, &initValue, sizeof(initValue), allocSize, nullptr, 0, nullptr); - - ze_group_size_t groupSizes{groupSize, 1, 1}; - - size_t localWorkSizeForUint = groupSizes.groupSizeX * 4u; - - const void *kernelArgs[] = { - &dstBuffer, // output buffer - &localWorkSizeForUint, // local buffer for local ids - &localWorkSizeForUint // local buffer for global ids - }; - - SUCCESS_OR_TERMINATE(LevelZeroBlackBoxTests::zeCommandListAppendLaunchKernelWithArgumentsFunc(cmdList, kernel, groupCounts, groupSizes, kernelArgs, nullptr, nullptr, 0, nullptr)); - - SUCCESS_OR_TERMINATE(LevelZeroBlackBoxTests::zeDeviceSynchronizeFunc(device)); - - // Validate - outputValidationSuccessful = true; - - std::vector expectedOutput(groupCounts.groupCountX * 2, 0); - - for (auto i = 0; i < static_cast(groupCounts.groupCountX); ++i) { - auto sumOfLocalIds = groupSize * (groupSize - 1) / 2; // Sum of local IDs from 0 to localWorkSize-1 - auto maxGlobalId = groupSize * (i + 1) - 1; // max global id for this group - auto minGlobalId = groupSize * i; // min global id for this group - auto sumOfGlobalIdWithinGroup = (maxGlobalId * (maxGlobalId + 1) / 2) - (minGlobalId * (minGlobalId - 1) / 2); // sum of global ids within this group - - expectedOutput[i * 2] = sumOfLocalIds; - expectedOutput[i * 2 + 1] = sumOfGlobalIdWithinGroup; - } - for (auto i = 0; i < static_cast(expectedOutput.size()); ++i) { - auto expectedValue = expectedOutput[i]; - auto actualValue = reinterpret_cast(dstBuffer)[i]; - if (actualValue != expectedValue) { - std::cout << "dstBuffer[" << i << "] = " - << std::dec << actualValue << " not equal to " - << expectedValue << "\n"; - outputValidationSuccessful = false; - break; - } - } - if (!outputValidationSuccessful) { - break; - } - } - - // Cleanup - SUCCESS_OR_TERMINATE(zeMemFree(context, dstBuffer)); - SUCCESS_OR_TERMINATE(zeCommandListDestroy(cmdList)); - SUCCESS_OR_TERMINATE(zeKernelDestroy(kernel)); - SUCCESS_OR_TERMINATE(zeModuleDestroy(module)); -} - -int main(int argc, char *argv[]) { - LevelZeroBlackBoxTests::verbose = LevelZeroBlackBoxTests::isVerbose(argc, argv); - bool aubMode = LevelZeroBlackBoxTests::isAubMode(argc, argv); - - ze_context_handle_t context = nullptr; - const char *errorMsg = nullptr; - auto devices = LevelZeroBlackBoxTests::zelloInitContextAndGetDevices(context); - SUCCESS_OR_TERMINATE(LevelZeroBlackBoxTests::zerDriverGetLastErrorDescriptionFunc(&errorMsg)); - - if (errorMsg != nullptr && errorMsg[0] != 0) { - std::cerr << "Error initializing context: " << (errorMsg ? errorMsg : "Unknown error") << std::endl; - return 1; - } - - uint32_t deviceOrdinal = LevelZeroBlackBoxTests::zerDeviceTranslateToIdentifierFunc(devices[0]); - - SUCCESS_OR_TERMINATE(LevelZeroBlackBoxTests::zerDriverGetLastErrorDescriptionFunc(&errorMsg)); - - if (errorMsg != nullptr && errorMsg[0] != 0) { - std::cerr << "Error zerDeviceTranslateToIdentifier: " << errorMsg << std::endl; - return 1; - } - bool outputValidationSuccessful = false; - executeKernelAndValidate(context, deviceOrdinal, outputValidationSuccessful); - - LevelZeroBlackBoxTests::printResult(aubMode, outputValidationSuccessful, blackBoxName); - outputValidationSuccessful = aubMode ? true : outputValidationSuccessful; - return (outputValidationSuccessful ? 0 : 1); -} diff --git a/level_zero/core/test/unit_tests/sources/device/test_l0_device.cpp b/level_zero/core/test/unit_tests/sources/device/test_l0_device.cpp index f384cbbf04..b2d4243705 100644 --- a/level_zero/core/test/unit_tests/sources/device/test_l0_device.cpp +++ b/level_zero/core/test/unit_tests/sources/device/test_l0_device.cpp @@ -6521,7 +6521,6 @@ TEST_F(DeviceSimpleTests, whenWorkgroupSizeCheckedThenSizeLimitIs1kOrLess) { HWTEST_F(DeviceSimpleTests, givenGpuHangWhenSynchronizingDeviceThenErrorIsPropagated) { auto &csr = neoDevice->getUltCommandStreamReceiver(); - csr.resourcesInitialized = true; csr.waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::gpuHang; auto result = zeDeviceSynchronize(device); @@ -6541,7 +6540,6 @@ HWTEST_F(DeviceSimpleTests, givenNoGpuHangWhenSynchronizingDeviceThenCallWaitFor csr->flushStamp->setStamp(flushStampToWait++); csr->captureWaitForTaskCountWithKmdNotifyInputParams = true; csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::ready; - csr->resourcesInitialized = true; } auto &secondaryCsrs = neoDevice->getSecondaryCsrs(); @@ -6553,7 +6551,6 @@ HWTEST_F(DeviceSimpleTests, givenNoGpuHangWhenSynchronizingDeviceThenCallWaitFor csr->flushStamp->setStamp(flushStampToWait++); csr->captureWaitForTaskCountWithKmdNotifyInputParams = true; csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::ready; - csr->resourcesInitialized = true; } auto result = zeDeviceSynchronize(device); @@ -6578,48 +6575,6 @@ HWTEST_F(DeviceSimpleTests, givenNoGpuHangWhenSynchronizingDeviceThenCallWaitFor } } -HWTEST_F(DeviceSimpleTests, whenSynchronizingDeviceThenIgnoreUninitializedCsrs) { - auto &engines = neoDevice->getAllEngines(); - - TaskCountType taskCountToWait = 1u; - FlushStamp flushStampToWait = 4u; - for (auto &engine : engines) { - auto csr = static_cast *>(engine.commandStreamReceiver); - csr->latestSentTaskCount = 0u; - csr->latestFlushedTaskCount = 0u; - csr->taskCount = taskCountToWait++; - csr->flushStamp->setStamp(flushStampToWait++); - csr->captureWaitForTaskCountWithKmdNotifyInputParams = true; - csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::gpuHang; - csr->resourcesInitialized = false; - } - - auto &secondaryCsrs = neoDevice->getSecondaryCsrs(); - for (auto &secondaryCsr : secondaryCsrs) { - auto csr = static_cast *>(secondaryCsr.get()); - csr->latestSentTaskCount = 0u; - csr->latestFlushedTaskCount = 0u; - csr->taskCount = taskCountToWait++; - csr->flushStamp->setStamp(flushStampToWait++); - csr->captureWaitForTaskCountWithKmdNotifyInputParams = true; - csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::gpuHang; - csr->resourcesInitialized = false; - } - - auto result = zeDeviceSynchronize(device); - EXPECT_EQ(ZE_RESULT_SUCCESS, result); - - for (auto &engine : engines) { - auto csr = static_cast *>(engine.commandStreamReceiver); - EXPECT_EQ(0u, csr->waitForTaskCountWithKmdNotifyInputParams.size()); - } - - for (auto &secondaryCsr : secondaryCsrs) { - auto csr = static_cast *>(secondaryCsr.get()); - EXPECT_EQ(0u, csr->waitForTaskCountWithKmdNotifyInputParams.size()); - } -} - HWTEST_F(DeviceSimpleTests, givenGpuHangOnSecondaryCsrWhenSynchronizingDeviceThenErrorIsPropagated) { if (neoDevice->getSecondaryCsrs().empty()) { GTEST_SKIP(); @@ -6628,14 +6583,12 @@ HWTEST_F(DeviceSimpleTests, givenGpuHangOnSecondaryCsrWhenSynchronizingDeviceThe for (auto &engine : engines) { auto csr = static_cast *>(engine.commandStreamReceiver); - csr->resourcesInitialized = true; csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::ready; } auto &secondaryCsrs = neoDevice->getSecondaryCsrs(); for (auto &secondaryCsr : secondaryCsrs) { auto csr = static_cast *>(secondaryCsr.get()); - csr->resourcesInitialized = true; csr->waitForTaskCountWithKmdNotifyFallbackReturnValue = WaitStatus::gpuHang; }