From 38cfac4dcd8cf666fd852c2221ac4bab72e3029d Mon Sep 17 00:00:00 2001
From: foreman <dl.swbuild@amd.com>
Date: Wed, 24 Jul 2019 12:05:47 -0400
Subject: [PATCH] P4 to Git Change 1972922 by cpaquot@cpaquot-ocl-lc-lnx on
 2019/07/24 11:59:16

	SWDEV-196881 - [HIP] Fix hipStreamQuery implementation. Using isEmpty isn't correct.

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_internal.hpp#32 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_stream.cpp#20 edit


[ROCm/clr commit: 47ac573d70bc09e593efb55d2631fb48d10aba3a]
---
 projects/clr/hipamd/api/hip/hip_internal.hpp |  1 +
 projects/clr/hipamd/api/hip/hip_stream.cpp   | 13 ++++++++++++-
 2 files changed, 13 insertions(+), 1 deletion(-)

diff --git a/projects/clr/hipamd/api/hip/hip_internal.hpp b/projects/clr/hipamd/api/hip/hip_internal.hpp
index eae35ecae9..5d32f1049f 100644
--- a/projects/clr/hipamd/api/hip/hip_internal.hpp
+++ b/projects/clr/hipamd/api/hip/hip_internal.hpp
@@ -159,6 +159,7 @@ extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset);
 
 #define HIP_RETURN(ret)          \
         hip::g_lastError = ret;  \
+        LogPrintfInfo("[%zx] %s: Returned %s", std::this_thread::get_id(), __func__, hipGetErrorName(ret)); \
         return hip::g_lastError; \
 
 inline std::ostream& operator<<(std::ostream& os, const dim3& s) {
diff --git a/projects/clr/hipamd/api/hip/hip_stream.cpp b/projects/clr/hipamd/api/hip/hip_stream.cpp
index 59456303fe..cbba339efb 100644
--- a/projects/clr/hipamd/api/hip/hip_stream.cpp
+++ b/projects/clr/hipamd/api/hip/hip_stream.cpp
@@ -210,7 +210,18 @@ hipError_t hipStreamQuery(hipStream_t stream) {
   } else {
     hostQueue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
   }
-  HIP_RETURN(hostQueue->isEmpty() ? hipSuccess : hipErrorNotReady);
+
+  amd::Command* command = hostQueue->getLastQueuedCommand(false);
+  if (command == nullptr) {
+    HIP_RETURN(hipSuccess);
+  }
+
+  amd::Event& event = command->event();
+
+  if (command->type() != CL_COMMAND_MARKER) {
+    event.notifyCmdQueue();
+  }
+  HIP_RETURN((command->status() == CL_COMPLETE) ? hipSuccess : hipErrorNotReady);
 }
 
 hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback, void* userData,