From bb27a1d7e8624700dca331a66daa277119eff4a9 Mon Sep 17 00:00:00 2001 From: Ben Sander Date: Tue, 23 Feb 2016 04:07:05 -0600 Subject: [PATCH] Sync review. - add calls to ihipInit missing from some routines. - sync before draining a stream. [ROCm/clr commit: 8f98aca124794e3f71e8f9b1c7ea9ee88ba86321] --- projects/clr/hipamd/src/hip_hcc.cpp | 14 +++++++++++++- 1 file changed, 13 insertions(+), 1 deletion(-) diff --git a/projects/clr/hipamd/src/hip_hcc.cpp b/projects/clr/hipamd/src/hip_hcc.cpp index 3ba578d52c..f01579ae71 100644 --- a/projects/clr/hipamd/src/hip_hcc.cpp +++ b/projects/clr/hipamd/src/hip_hcc.cpp @@ -65,7 +65,6 @@ THE SOFTWARE. //static const int debug = 0; static const int release = 1; -#define ENABLE_CHECKS 1 int HIP_LAUNCH_BLOCKING = 0; @@ -1413,6 +1412,15 @@ hipError_t hipStreamDestroy(hipStream_t stream) hipError_t e = hipSuccess; + //--- Drain the stream: + if (stream == NULL) { + ihipDevice_t *device = ihipGetTlsDefaultDevice(); + ihipWaitNullStream(device); + } else { + stream->wait(); + e = hipSuccess; + } + ihipDevice_t *device = stream->getDevice(); if (device) { @@ -1862,6 +1870,8 @@ hipError_t hipMallocHost(void** ptr, size_t sizeBytes) //--- hipError_t hipMemcpyToSymbol(const char* symbolName, const void *src, size_t count, size_t offset, hipMemcpyKind kind) { + std::call_once(hip_initialized, ihipInit); + #ifdef USE_MEMCPYTOSYMBOL if(kind != hipMemcpyHostToDevice) { @@ -2288,6 +2298,8 @@ hipError_t hipMemsetAsync(void* dst, int value, size_t sizeBytes, hipStream_t s hipError_t hipMemset(void* dst, int value, size_t sizeBytes ) { + std::call_once(hip_initialized, ihipInit); + // TODO - call an ihip memset so HIP_TRACE is correct. return hipMemsetAsync(dst, value, sizeBytes, hipStreamNull); }