Fixed a few multithreaded potential issues

Also make D2H and H2D keep track of the chain of events
when we need to use a different HostQueue.

Change-Id: I1c5da6ea6104b37ad7aac00f0eb8ea9371e6ba1c

hipamd/vdi/hip_memory.cpp

@@ -102,6 +102,7 @@ hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKin
     amd::HostQueue* pQueue = &queue;
     if (queueDevice != dstMemory->getContext().devices()[0]) {
       pQueue = hip::getNullStream(dstMemory->getContext());
+      waitList.push_back(queue.getLastQueuedCommand(true));
     }
     command = new amd::WriteMemoryCommand(*pQueue, CL_COMMAND_WRITE_BUFFER, waitList,
               *dstMemory->asBuffer(), dOffset, sizeBytes, src);
@@ -110,6 +111,7 @@ hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKin
     amd::HostQueue* pQueue = &queue;
     if (queueDevice != srcMemory->getContext().devices()[0]) {
       pQueue = hip::getNullStream(srcMemory->getContext());
+      waitList.push_back(queue.getLastQueuedCommand(true));
     }
     command = new amd::ReadMemoryCommand(*pQueue, CL_COMMAND_READ_BUFFER, waitList,
               *srcMemory->asBuffer(), sOffset, sizeBytes, dst);
@@ -155,6 +157,10 @@ hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKin
   }
   command->release();
 
+  if (waitList.size() > 0) {
+    waitList[0]->release();
+  }
+
   return hipSuccess;
 }
 

hipamd/vdi/hip_stream.cpp

@@ -224,7 +224,7 @@ hipError_t hipStreamQuery(hipStream_t stream) {
     hostQueue = reinterpret_cast<hip::Stream*>(stream)->asHostQueue();
   }
 
-  amd::Command* command = hostQueue->getLastQueuedCommand(false);
+  amd::Command* command = hostQueue->getLastQueuedCommand(true);
   if (command == nullptr) {
     HIP_RETURN(hipSuccess);
   }
@@ -233,7 +233,9 @@ hipError_t hipStreamQuery(hipStream_t stream) {
   if (command->type() != 0) {
     event.notifyCmdQueue();
   }
-  HIP_RETURN((command->status() == CL_COMPLETE) ? hipSuccess : hipErrorNotReady);
+  hipError_t status = (command->status() == CL_COMPLETE) ? hipSuccess : hipErrorNotReady;
+  command->release();
+  HIP_RETURN(status);
 }
 
 hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback, void* userData,