From 6bb83fb95aa75dca6457466937d63cedfbf7353b Mon Sep 17 00:00:00 2001 From: "Mrozek, Michal" Date: Fri, 9 Feb 2018 11:57:44 +0100 Subject: [PATCH] Do not noop pipe controls if call is returning event on IOQ. -For in order queue application can have fine grain granularity of completion -For out of order queue application wants to execute workloads concurrently -This change disables pipe control nooping for ioq calls when event returned. Change-Id: Iaeaf677f768f7434b2efa1842b50653ab80777ad --- runtime/command_queue/enqueue_common.h | 2 +- .../command_queue/enqueue_kernel_tests.cpp | 50 +++++++++++++++++++ 2 files changed, 51 insertions(+), 1 deletion(-) diff --git a/runtime/command_queue/enqueue_common.h b/runtime/command_queue/enqueue_common.h index 089f23dd8f..6d54388610 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 = true; + dispatchFlags.outOfOrderExecutionAllowed = !eventBuilder.getEvent() || this->isOOQEnabled(); 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 e35b812b8e..1f63a5250b 100644 --- a/unit_tests/command_queue/enqueue_kernel_tests.cpp +++ b/unit_tests/command_queue/enqueue_kernel_tests.cpp @@ -1443,6 +1443,56 @@ HWTEST_F(EnqueueKernelTest, givenInOrderCommandQueueWhenEnqueueKernelIsMadeThenP clReleaseCommandQueue(inOrderQueue); } +HWTEST_F(EnqueueKernelTest, givenInOrderCommandQueueWhenEnqueueKernelReturningEventIsMadeThenPipeControlPositionIsRecorded) { + 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}; + cl_event event; + + clEnqueueNDRangeKernel(inOrderQueue, mockKernel.mockKernel, 1, nullptr, gws, nullptr, 0, nullptr, &event); + + EXPECT_FALSE(mockedSubmissionsAggregator->peekCmdBufferList().peekIsEmpty()); + auto cmdBuffer = mockedSubmissionsAggregator->peekCmdBufferList().peekHead(); + EXPECT_EQ(nullptr, cmdBuffer->pipeControlLocation); + + clReleaseCommandQueue(inOrderQueue); + clReleaseEvent(event); +} + +HWTEST_F(EnqueueKernelTest, givenOutOfOrderCommandQueueWhenEnqueueKernelReturningEventIsMadeThenPipeControlPositionIsRecorded) { + const cl_queue_properties props[3] = {CL_QUEUE_PROPERTIES, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, 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}; + cl_event event; + + clEnqueueNDRangeKernel(inOrderQueue, mockKernel.mockKernel, 1, nullptr, gws, nullptr, 0, nullptr, &event); + + EXPECT_FALSE(mockedSubmissionsAggregator->peekCmdBufferList().peekIsEmpty()); + auto cmdBuffer = mockedSubmissionsAggregator->peekCmdBufferList().peekHead(); + EXPECT_NE(nullptr, cmdBuffer->pipeControlLocation); + + clReleaseCommandQueue(inOrderQueue); + clReleaseEvent(event); +} + HWTEST_F(EnqueueKernelTest, givenCsrInBatchingModeWhenBlockingCallIsMadeThenEventAssociatedWithCommandHasProperFlushStamp) { DebugManagerStateRestore stateRestore; DebugManager.flags.MakeEachEnqueueBlocking.set(true);