From 39ea36a4589d6c06be3994e1ec4e28c201e6bd69 Mon Sep 17 00:00:00 2001
From: foreman <dl.constructicon@amd.com>
Date: Wed, 9 May 2018 18:15:41 -0400
Subject: [PATCH] P4 to Git Change 1552807 by
 skudchad@skudchad_test2_win_opencl on 2018/05/09 18:08:07

	SWDEV-145570 - [HIP] - Release a stream first before taking it off from the set.
	- Queue::create() needs to be called before returning a valid queue.

	ReviewBoardURL = http://ocltc.amd.com/reviews/r/14830/diff/

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_stream.cpp#7 edit


[ROCm/clr commit: 1426982a6560f4a1382e5521559a1fcf3f532b0e]
---
 projects/clr/hipamd/api/hip/hip_stream.cpp | 12 +++++-------
 1 file changed, 5 insertions(+), 7 deletions(-)

diff --git a/projects/clr/hipamd/api/hip/hip_stream.cpp b/projects/clr/hipamd/api/hip/hip_stream.cpp
index 1c007edf1d..7cb359b1b9 100644
--- a/projects/clr/hipamd/api/hip/hip_stream.cpp
+++ b/projects/clr/hipamd/api/hip/hip_stream.cpp
@@ -32,15 +32,15 @@ static hipError_t ihipStreamCreateWithFlags(hipStream_t *stream, unsigned int fl
                                              amd::CommandQueue::RealTimeDisabled,
                                              amd::CommandQueue::Priority::Normal);
 
+  if (queue == nullptr || !queue->create()) {
+    return hipErrorOutOfMemory;
+  }
+
   if (!(flags & hipStreamNonBlocking)) {
     hip::syncStreams();
     streamSet.insert(queue);
   }
 
-  if (queue == nullptr) {
-    return hipErrorOutOfMemory;
-  }
-
   *stream = reinterpret_cast<hipStream_t>(as_cl(queue));
 
   return hipSuccess;
@@ -101,10 +101,8 @@ hipError_t hipStreamDestroy(hipStream_t stream) {
   }
 
   amd::HostQueue* hostQueue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
-
-  streamSet.erase(hostQueue);
-
   hostQueue->release();
+  streamSet.erase(hostQueue);
 
   return hipSuccess;
 }