diff --git a/src/hip_hcc.cpp b/src/hip_hcc.cpp index 4edf9575be..c0a3827376 100644 --- a/src/hip_hcc.cpp +++ b/src/hip_hcc.cpp @@ -1038,7 +1038,7 @@ void ihipCtx_t::locked_syncDefaultStream(bool waitOnSelf, bool syncHost) { } } else { if (waitThisStream) { - LockedAccessor_StreamCrit_t streamCrit(stream->_criticalData); + LockedAccessor_StreamCrit_t streamCrit(stream->criticalData()); // The last marker will provide appropriate visibility: if (!streamCrit->_av.get_is_empty()) { @@ -1056,7 +1056,7 @@ void ihipCtx_t::locked_syncDefaultStream(bool waitOnSelf, bool syncHost) { // Enqueue a barrier to wait on all the barriers we sent above: if (!HIP_SYNC_NULL_STREAM && !depOps.empty()) { - LockedAccessor_StreamCrit_t defaultStreamCrit(_defaultStream->_criticalData); + LockedAccessor_StreamCrit_t defaultStreamCrit(_defaultStream->criticalData()); tprintf(DB_SYNC, " null-stream wait on %zu non-empty streams. sync_host=%d\n", depOps.size(), syncHost); hc::completion_future defaultCf = defaultStreamCrit->_av.create_blocking_marker( diff --git a/src/hip_hcc_internal.h b/src/hip_hcc_internal.h index 852e8985ad..a07ff0e7e7 100644 --- a/src/hip_hcc_internal.h +++ b/src/hip_hcc_internal.h @@ -609,12 +609,10 @@ class ihipStream_t { void addSymbolPtrToTracker(hc::accelerator& acc, void* ptr, size_t sizeBytes); - - public: // TODO - move private + private: // Data // Critical Data - MUST be accessed through LockedAccessor_StreamCrit_t ihipStreamCritical_t _criticalData; - private: // Data std::mutex _hasQueueLock; ihipCtx_t* _ctx; // parent context that owns this stream. diff --git a/src/hip_stream.cpp b/src/hip_stream.cpp index b748749538..d70f860f90 100644 --- a/src/hip_stream.cpp +++ b/src/hip_stream.cpp @@ -246,7 +246,7 @@ hipError_t hipStreamGetPriority(hipStream_t stream, int* priority) { #if defined(__HCC__) && (__hcc_minor__ < 3) *priority = 0; #else - LockedAccessor_StreamCrit_t crit(stream->_criticalData); + LockedAccessor_StreamCrit_t crit(stream->criticalData()); *priority = crit->_av.get_queue_priority(); #endif return ihipLogStatus(hipSuccess);