SWDEV-514950 - [6.4 Preview] fix hipMallocManaged and hipMemAdvise during stream capture (#128)

lttamd · web-flow · commit 9b2f84729a85 · 2025-05-14T21:40:05.000+05:30
diff --git a/hipamd/src/hip_hmm.cpp b/hipamd/src/hip_hmm.cpp
@@ -72,6 +72,10 @@ hipError_t hipMallocManaged(void** dev_ptr, size_t size, unsigned int flags) {
     HIP_RETURN(hipErrorInvalidValue);
   }
 
+  if (!hip::tls.capture_streams_.empty() || !g_captureStreams.empty()) {
+    HIP_RETURN(hipErrorStreamCaptureUnsupported);
+  }
+
   HIP_RETURN(ihipMallocManaged(dev_ptr, size), *dev_ptr);
 }
 
@@ -153,6 +157,10 @@ hipError_t hipMemAdvise(const void* dev_ptr, size_t count, hipMemoryAdvise advic
     HIP_RETURN(hipErrorInvalidValue);
   }
 
+  if (!hip::tls.capture_streams_.empty() || !g_captureStreams.empty()) {
+    HIP_RETURN(hipErrorStreamCaptureUnsupported);
+  }
+
   size_t offset = 0;
   amd::Memory* memObj = getMemoryObject(dev_ptr, offset);
   if (memObj && count > (memObj->getSize() - offset)) {

Original file line number	Diff line number	Diff line change
`@@ -72,6 +72,10 @@ hipError_t hipMallocManaged(void** dev_ptr, size_t size, unsigned int flags) {`
`72`	`72`	`HIP_RETURN(hipErrorInvalidValue);`
`73`	`73`	`}`
`74`	`74`
	`75`	`+ if (!hip::tls.capture_streams_.empty() \|\| !g_captureStreams.empty()) {`
	`76`	`+ HIP_RETURN(hipErrorStreamCaptureUnsupported);`
	`77`	`+ }`
	`78`	`+`
`75`	`79`	`HIP_RETURN(ihipMallocManaged(dev_ptr, size), *dev_ptr);`
`76`	`80`	`}`
`77`	`81`
`@@ -153,6 +157,10 @@ hipError_t hipMemAdvise(const void* dev_ptr, size_t count, hipMemoryAdvise advic`
`153`	`157`	`HIP_RETURN(hipErrorInvalidValue);`
`154`	`158`	`}`
`155`	`159`
	`160`	`+ if (!hip::tls.capture_streams_.empty() \|\| !g_captureStreams.empty()) {`
	`161`	`+ HIP_RETURN(hipErrorStreamCaptureUnsupported);`
	`162`	`+ }`
	`163`	`+`
`156`	`164`	`size_t offset = 0;`
`157`	`165`	`amd::Memory* memObj = getMemoryObject(dev_ptr, offset);`
`158`	`166`	`if (memObj && count > (memObj->getSize() - offset)) {`