From 56154e33dcf4fcdf9da2024f5e3e7ebcd521c8c2 Mon Sep 17 00:00:00 2001 From: German Andryeyev Date: Mon, 22 Feb 2021 23:22:44 -0500 Subject: [PATCH] SWDEV-272496 - Add marker if notification was the last command. If MT is enabled, then a new callback can be received before the previous command is processed, causing a conflict of 2 callbacks. Change-Id: I5ff8f231208e8d62824d590d3c8e791e8e36affb --- rocclr/hip_stream.cpp | 6 +++++- 1 file changed, 5 insertions(+), 1 deletion(-) diff --git a/rocclr/hip_stream.cpp b/rocclr/hip_stream.cpp index 604f3ea33d..01820476cb 100755 --- a/rocclr/hip_stream.cpp +++ b/rocclr/hip_stream.cpp @@ -368,7 +368,11 @@ hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback } amd::HostQueue* hostQueue = hip::getQueue(stream); amd::Command* command = hostQueue->getLastQueuedCommand(true); - if (command == nullptr) { + if (command == nullptr || + // Add a marker if the last command is a notification, + // since with multithreading AddCallback may occur before the previous + // command can be processed + (command->type() == 0)) { amd::Command::EventWaitList eventWaitList; command = new amd::Marker(*hostQueue, kMarkerDisableFlush, eventWaitList); command->enqueue();