diff --git a/rocclr/device/rocm/rocmemory.cpp b/rocclr/device/rocm/rocmemory.cpp
index 9887d23334..763317df22 100644
--- a/rocclr/device/rocm/rocmemory.cpp
+++ b/rocclr/device/rocm/rocmemory.cpp
@@ -615,7 +615,7 @@ Buffer::Buffer(const roc::Device& dev, size_t size) : roc::Memory(dev, size) {}
 
 Buffer::~Buffer() {
   if (owner() == nullptr) {
-    dev().hostFree(deviceMemory_, size());
+    dev().memFree(deviceMemory_, size());
   } else {
     destroy();
 
@@ -657,7 +657,7 @@ void Buffer::destroy() {
               ClPrint(amd::LOG_DEBUG, amd::LOG_MEM, "[ROCClr] munmap failed \n");
             }
           } else {
-            dev().hostFree(deviceMemory_, size());
+            dev().memFree(deviceMemory_, size());
           }
         } else if (memFlags & ROCCLR_MEM_HSA_SIGNAL_MEMORY) {
           if (HSA_STATUS_SUCCESS != hsa_signal_destroy(signal_)) {
@@ -666,7 +666,7 @@ void Buffer::destroy() {
           }
           deviceMemory_ = nullptr;
         } else {
-          dev().hostFree(deviceMemory_, size());
+          dev().memFree(deviceMemory_, size());
         }
       } else {
         dev().memFree(deviceMemory_, size());
diff --git a/rocclr/platform/runtime.cpp b/rocclr/platform/runtime.cpp
index 7f80919c46..6072a62a9b 100644
--- a/rocclr/platform/runtime.cpp
+++ b/rocclr/platform/runtime.cpp
@@ -120,4 +120,12 @@ uint ReferenceCountedObject::release() {
   return newCount;
 }
 
+#ifndef _WIN32
+void __attribute__((destructor)) hipTearDown() {
+  if (amd::IS_HIP) {
+    Runtime::tearDown();
+  }
+}
+#endif
+
 }  // namespace amd