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

@@ -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();