SWDEV-444670 - Enable teardown class

Force implicit runtime teardown with a global destructor.

Change-Id: Iabe63dedf5b94fefc98668585c45a61607120669


[ROCm/clr commit: c95a75a2bf]

projects/clr/hipamd/src/hip_context.cpp

@@ -61,6 +61,7 @@ void init(bool* status) {
       return;
     }
     g_devices.push_back(device);
+    amd::RuntimeTearDown::RegisterObject(device);
   }
 
   amd::Context* hContext = new amd::Context(devices, amd::Context::Info());
@@ -73,6 +74,7 @@ void init(bool* status) {
     hContext->release();
   }
   host_context = hContext;
+  amd::RuntimeTearDown::RegisterObject(hContext);
 
   PlatformState::instance().init();
   *status = true;

projects/clr/hipamd/src/hip_internal.hpp

@@ -445,7 +445,7 @@ public:
   };
 
   /// HIP Device class
-  class Device {
+  class Device : public amd::ReferenceCountedObject {
     amd::Monitor lock_{"Device lock", true};
     amd::Monitor streamSetLock{"Guards device stream set"};
     std::unordered_set<hip::Stream*> streamSet;

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

@@ -259,7 +259,7 @@ Device::~Device() {
     p2p_stage_ = nullptr;
   }
   if (nullptr != mg_sync_) {
-    amd::SvmBuffer::free(GlbCtx(), mg_sync_);
+    GlbCtx().svmFree(mg_sync_);
     mg_sync_ = nullptr;
   }
   if (glb_ctx_ != nullptr) {
@@ -619,9 +619,8 @@ bool Device::init() {
 
     // Allocate mgpu sync buffer for cooperative launches
     if (amd::IS_HIP) {
-      mg_sync_ = reinterpret_cast<address>(amd::SvmBuffer::malloc(
-          *glb_ctx_, (CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS),
-          kMGInfoSizePerDevice * devices.size(), kMGInfoSizePerDevice));
+      mg_sync_ = reinterpret_cast<address>(glb_ctx_->svmAlloc(kMGInfoSizePerDevice * devices.size(),
+        kMGInfoSizePerDevice, (CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS)));
       if (mg_sync_ == nullptr) {
         LogError("mgpu sync buffer alloc failed");
         return false;

projects/clr/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();
 
@@ -663,7 +663,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_)) {
@@ -672,7 +672,7 @@ void Buffer::destroy() {
           }
           deviceMemory_ = nullptr;
         } else {
-          dev().hostFree(deviceMemory_, size());
+          dev().memFree(deviceMemory_, size());
         }
       } else {
         dev().memFree(deviceMemory_, size());

projects/clr/rocclr/platform/memory.cpp

@@ -1521,7 +1521,6 @@ bool SvmBuffer::Contains(uintptr_t ptr) {
 // The allocation flags are ignored for now.
 void* SvmBuffer::malloc(Context& context, cl_svm_mem_flags flags, size_t size, size_t alignment,
                         const amd::Device* curDev) {
-  bool atomics = (flags & CL_MEM_SVM_ATOMICS) != 0;
   void* ret = context.svmAlloc(size, alignment, flags, curDev);
   if (ret == nullptr) {
     LogError("Unable to allocate aligned memory");

projects/clr/rocclr/platform/runtime.cpp

@@ -100,11 +100,24 @@ void Runtime::tearDown() {
   initialized_ = false;
 }
 
-class RuntimeTearDown : public amd::HeapObject {
-public:
-  RuntimeTearDown() {}
-  ~RuntimeTearDown() { /*Runtime::tearDown();*/ }
-} runtime_tear_down;
+std::vector<ReferenceCountedObject*> RuntimeTearDown::external_;
+
+RuntimeTearDown::~RuntimeTearDown() {
+#ifndef _WIN32
+  if (amd::IS_HIP) {
+    for (auto it: external_) {
+      it->release();
+    }
+    Runtime::tearDown();
+  }
+#endif
+}
+
+void RuntimeTearDown::RegisterObject(ReferenceCountedObject* obj) {
+    external_.push_back(obj);
+}
+
+class RuntimeTearDown runtime_tear_down;
 
 uint ReferenceCountedObject::retain() {
   return referenceCount_.fetch_add(1, std::memory_order_relaxed) + 1;

projects/clr/rocclr/platform/runtime.hpp

@@ -91,6 +91,16 @@ public:
 
 inline bool Runtime::initialized() { return initialized_; }
 
+class RuntimeTearDown : public HeapObject {
+  static std::vector<ReferenceCountedObject*> external_;
+
+public:
+  RuntimeTearDown() {}
+  ~RuntimeTearDown();
+
+  static void RegisterObject(ReferenceCountedObject* obj);
+};
+
 }  // namespace amd
 
 #endif /*RUNTIME_HPP_*/