diff --git a/runtime/command_queue/enqueue_common.h b/runtime/command_queue/enqueue_common.h index ad24ab4cf8..089f23dd8f 100644 --- a/runtime/command_queue/enqueue_common.h +++ b/runtime/command_queue/enqueue_common.h @@ -538,7 +538,7 @@ CompletionStamp CommandQueueHw::enqueueNonBlocked( dispatchFlags.implicitFlush = implicitFlush; dispatchFlags.flushStampReference = this->flushStamp->getStampReference(); dispatchFlags.preemptionMode = PreemptionHelper::taskPreemptionMode(*device, multiDispatchInfo); - dispatchFlags.outOfOrderExecutionAllowed = this->isOOQEnabled(); + dispatchFlags.outOfOrderExecutionAllowed = true; DEBUG_BREAK_IF(taskLevel >= Event::eventNotReady); diff --git a/unit_tests/command_queue/enqueue_kernel_tests.cpp b/unit_tests/command_queue/enqueue_kernel_tests.cpp index 49d544249b..7feeb6ce66 100644 --- a/unit_tests/command_queue/enqueue_kernel_tests.cpp +++ b/unit_tests/command_queue/enqueue_kernel_tests.cpp @@ -1421,6 +1421,28 @@ HWTEST_F(EnqueueKernelTest, givenOutOfOrderCommandQueueWhenEnqueueKernelIsMadeTh clReleaseCommandQueue(ooq); } +HWTEST_F(EnqueueKernelTest, givenInOrderCommandQueueWhenEnqueueKernelIsMadeThenPipeControlPositionIsRecorded) { + const cl_queue_properties props[] = {0}; + auto inOrderQueue = clCreateCommandQueueWithProperties(context, pDevice, props, nullptr); + + auto mockCsr = new MockCsrHw2(pDevice->getHardwareInfo()); + mockCsr->overrideDispatchPolicy(CommandStreamReceiver::DispatchMode::BatchedDispatch); + pDevice->resetCommandStreamReceiver(mockCsr); + + auto mockedSubmissionsAggregator = new mockSubmissionsAggregator(); + mockCsr->overrideSubmissionAggregator(mockedSubmissionsAggregator); + + MockKernelWithInternals mockKernel(*pDevice); + size_t gws[3] = {1, 0, 0}; + clEnqueueNDRangeKernel(inOrderQueue, mockKernel.mockKernel, 1, nullptr, gws, nullptr, 0, nullptr, nullptr); + + EXPECT_FALSE(mockedSubmissionsAggregator->peekCmdBufferList().peekIsEmpty()); + auto cmdBuffer = mockedSubmissionsAggregator->peekCmdBufferList().peekHead(); + EXPECT_NE(nullptr, cmdBuffer->pipeControlLocation); + + clReleaseCommandQueue(inOrderQueue); +} + HWTEST_F(EnqueueKernelTest, givenCsrInBatchingModeWhenBlockingCallIsMadeThenEventAssociatedWithCommandHasProperFlushStamp) { DebugManagerStateRestore stateRestore; DebugManager.flags.MakeEachEnqueueBlocking.set(true);