Merge branch 'develop' into users/sluzynsk-amd/SWDEV-546322-3

Author: sluzynsk-amd

projects/clr/hipamd/include/hip/amd_detail/hip_prof_str.h

@@ -11610,7 +11610,7 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da
       oss << ", mode="; roctracer::hip_support::detail::operator<<(oss, data->args.hipStreamBeginCaptureToGraph.mode);
       oss << ")";
     break;
-     case HIP_API_ID_hipStreamCopyAttributes:
+    case HIP_API_ID_hipStreamCopyAttributes:
       oss << "hipStreamCopyAttributes(";
       oss << "dst="; roctracer::hip_support::detail::operator<<(oss, data->args.hipStreamCopyAttributes.dst);
       oss << ", src="; roctracer::hip_support::detail::operator<<(oss, data->args.hipStreamCopyAttributes.src);

projects/clr/rocclr/device/rocm/rocblit.cpp

@@ -729,6 +729,8 @@ bool DmaBlitManager::hsaCopyStagedOrPinned(const_address hostSrc, address hostDs
       address dst = hostDst + copyOffset;
       status = rocrCopyBuffer(dst, dstAgent, stagingBuffer, srcAgent, copysize, copyMetadata);
       if (!status) {
+        // Release Pinned Memory back to pool if any
+        releaseBuffer(outBuffer);
         break;
       }
     } else {  // D2H Path
@@ -745,6 +747,8 @@ bool DmaBlitManager::hsaCopyStagedOrPinned(const_address hostSrc, address hostDs
           memcpy(hostDst + copyOffset, stagingBuffer, copysize);
         }
       } else {
+        // Release Pinned Memory back to pool if any
+        releaseBuffer(outBuffer);
         break;
       }
     }