diff --git a/src/core/roctracer.cpp b/src/core/roctracer.cpp
index 819aa018f1..26e5ddbcc5 100644
--- a/src/core/roctracer.cpp
+++ b/src/core/roctracer.cpp
@@ -526,7 +526,7 @@ hsa_status_t hsa_amd_memory_async_copy_interceptor(
     status = hsa_amd_memory_async_copy_fn(dst, dst_agent, src,
                                           src_agent, size, num_dep_signals,
                                           dep_signals, entry->signal);
-    if (status != HSA_STATUS_SUCCESS) EXC_ABORT(status, "hsa_amd_memory_async_copy interceptor");
+    if (status != HSA_STATUS_SUCCESS) ::proxy::Tracker::Disable(entry);
   }
   else
   {
@@ -551,7 +551,7 @@ hsa_status_t hsa_amd_memory_async_copy_rect_interceptor(
                                                src_offset, range, copy_agent,
                                                dir, num_dep_signals, dep_signals,
                                                entry->signal);
-    if (status != HSA_STATUS_SUCCESS) EXC_ABORT(status, "hsa_amd_memory_async_copy interceptor");
+    if (status != HSA_STATUS_SUCCESS) ::proxy::Tracker::Disable(entry);
   }
   else
   {
@@ -985,36 +985,51 @@ PUBLIC_API roctracer_status_t roctracer_set_properties(
 // HSA-runtime tool on-load method
 PUBLIC_API bool roctracer_load(HsaApiTable* table, uint64_t runtime_version, uint64_t failed_tool_count,
                        const char* const* failed_tool_names) {
+//  printf("LIB roctracer_load\n"); fflush(stdout);
   static bool is_loaded = false;
   if (is_loaded) return true;
   is_loaded = true;
 
+//  printf("LIB roctracer_load end\n"); fflush(stdout);
   return true;
 }
 
 PUBLIC_API void roctracer_unload(bool destruct) {
+//  printf("LIB roctracer_unload\n"); fflush(stdout);
   static bool is_unloaded = false;
   if (is_unloaded) return;
   is_unloaded = true;
 
   if (destruct == false) roctracer::trace_buffer.Flush();
   if ((roctracer::hsa_support::output_prefix != NULL) && (roctracer::kernel_file_handle != NULL)) fclose(roctracer::kernel_file_handle);
+//  printf("LIB roctracer_unload end\n"); fflush(stdout);
 }
 
 PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, uint64_t failed_tool_count,
                        const char* const* failed_tool_names) {
-  return roctracer_load(table, runtime_version, failed_tool_count, failed_tool_names);
+//  printf("LIB OnLoad\n"); fflush(stdout);
+  const bool ret = roctracer_load(table, runtime_version, failed_tool_count, failed_tool_names);
+//  printf("LIB OnLoad end\n"); fflush(stdout);
+  return ret;
+}
+PUBLIC_API void OnUnload() {
+//  printf("LIB OnUnload\n"); fflush(stdout);
+  roctracer_unload(false);
+//  printf("LIB OnUnload end\n"); fflush(stdout);
 }
-PUBLIC_API void OnUnload() { roctracer_unload(false); }
 
 CONSTRUCTOR_API void constructor() {
+//  printf("LIB constructor\n"); fflush(stdout);
   roctracer::util::Logger::Create();
+//  printf("LIB constructor end\n"); fflush(stdout);
 }
 
 DESTRUCTOR_API void destructor() {
+//  printf("LIB destructor\n"); fflush(stdout);
   roctracer_unload(true);
   util::HsaRsrcFactory::Destroy();
   roctracer::util::Logger::Destroy();
+//  printf("LIB destructor end\n"); fflush(stdout);
 }
 
 }  // extern "C"
diff --git a/src/core/trace_buffer.h b/src/core/trace_buffer.h
index 7086c7dea7..2f987218a8 100644
--- a/src/core/trace_buffer.h
+++ b/src/core/trace_buffer.h
@@ -4,6 +4,7 @@
 #include <list>
 #include <mutex>
 #include <pthread.h>
+#include <string.h>
 
 #define PTHREAD_CALL(call)                                                                         \
   do {                                                                                             \
@@ -61,8 +62,10 @@ class TraceBuffer {
     callback_t fun;
   };
 
-  TraceBuffer(const char* name, uint32_t size, flush_prm_t* flush_prm_arr, uint32_t flush_prm_count) {
-    (void) name;
+  TraceBuffer(const char* name, uint32_t size, flush_prm_t* flush_prm_arr, uint32_t flush_prm_count) :
+    is_flushed_(ATOMIC_FLAG_INIT)
+  {
+    name_ = strdup(name);
     size_ = size;
     data_ = allocate_fun();
     next_ = NULL;
@@ -72,7 +75,6 @@ class TraceBuffer {
 
     flush_prm_arr_ = flush_prm_arr;
     flush_prm_count_ = flush_prm_count;
-    is_flushed_ = false;
 
     PTHREAD_CALL(pthread_mutex_init(&work_mutex_, NULL));
     PTHREAD_CALL(pthread_cond_init(&work_cond_, NULL));
@@ -85,7 +87,7 @@ class TraceBuffer {
     PTHREAD_CALL(pthread_join(work_thread_, &res));
     if (res != PTHREAD_CANCELED) abort_run("~TraceBuffer: consumer thread wasn't stopped correctly");
 
-    if (is_flushed_ == false) flush_buf();
+    Flush();
   }
 
 
@@ -96,28 +98,28 @@ class TraceBuffer {
   }
 
   void Flush() {
-    PTHREAD_CALL(pthread_mutex_lock(&work_mutex_));
     flush_buf();
-    PTHREAD_CALL(pthread_mutex_unlock(&work_mutex_));
   }
 
   private:
   void flush_buf() {
-    is_flushed_ = true;
-    for (flush_prm_t* prm = flush_prm_arr_; prm < flush_prm_arr_ + flush_prm_count_; prm++) {
-      uint32_t type = prm->type;
-      callback_t fun = prm->fun;
-      pointer_t pointer = 0;
-      for (Entry* ptr : buf_list_) {
-        Entry* end = ptr + size_;
-        while ((ptr < end) && (pointer < read_pointer_)) {
-          if (ptr->type == type) {
-            if (ptr->valid == TRACE_ENTRY_COMPL) {
-              fun(ptr);
+    const bool is_flushed = atomic_flag_test_and_set_explicit(&is_flushed_, std::memory_order_acquire);
+    if (is_flushed == false) {
+      for (flush_prm_t* prm = flush_prm_arr_; prm < flush_prm_arr_ + flush_prm_count_; prm++) {
+        uint32_t type = prm->type;
+        callback_t fun = prm->fun;
+        pointer_t pointer = 0;
+        for (Entry* ptr : buf_list_) {
+          Entry* end = ptr + size_;
+          while ((ptr < end) && (pointer < read_pointer_)) {
+            if (ptr->type == type) {
+              if (ptr->valid == TRACE_ENTRY_COMPL) {
+                fun(ptr);
+              }
             }
+            ptr++;
+            pointer++;
           }
-          ptr++;
-          pointer++;
         }
       }
     }
@@ -164,6 +166,7 @@ class TraceBuffer {
     abort();
   }
 
+  const char* name_;
   uint32_t size_;
   Entry* data_;
   Entry* next_;
@@ -173,7 +176,7 @@ class TraceBuffer {
 
   flush_prm_t* flush_prm_arr_;
   uint32_t flush_prm_count_;
-  bool is_flushed_;
+  volatile std::atomic_flag is_flushed_;
 
   pthread_t work_thread_;
   pthread_mutex_t work_mutex_;
diff --git a/src/proxy/tracker.h b/src/proxy/tracker.h
index 40b41438aa..267f811a2f 100644
--- a/src/proxy/tracker.h
+++ b/src/proxy/tracker.h
@@ -60,6 +60,12 @@ class Tracker {
     if (status != HSA_STATUS_SUCCESS) EXC_RAISING(status, "hsa_amd_signal_async_handler");
   }
 
+  // Delete tracker entry
+  inline static void Disable(entry_t* entry) {
+    hsa_signal_destroy(entry->signal);
+    entry->valid.store(roctracer::TRACE_ENTRY_INV, std::memory_order_release);
+  }
+
   private:
   // Entry completion
   inline static void Complete(hsa_signal_value_t signal_value, entry_t* entry) {
diff --git a/test/tool/tracer_tool.cpp b/test/tool/tracer_tool.cpp
index dd5b711a18..14e3dfc4e8 100644
--- a/test/tool/tracer_tool.cpp
+++ b/test/tool/tracer_tool.cpp
@@ -367,6 +367,7 @@ FILE* open_output_file(const char* prefix, const char* name) {
 // HSA-runtime tool on-load method
 extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, uint64_t failed_tool_count,
                        const char* const* failed_tool_names) {
+//  printf("TOOL OnLoad\n"); fflush(stdout);
   timer = new hsa_rt_utils::Timer(table->core_->hsa_system_get_info_fn);
 
   // API traces switches
@@ -497,11 +498,13 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version,
     roctracer_set_properties(ACTIVITY_DOMAIN_HIP_API, (void*)mark_api_callback);
   }
 
+//  printf("TOOL OnLoad end\n"); fflush(stdout);
   return roctracer_load(table, runtime_version, failed_tool_count, failed_tool_names);
 }
 
 // tool unload method
 void tool_unload(bool destruct) {
+//  printf("TOOL tool_unload\n"); fflush(stdout);
   static bool is_unloaded = false;
   if (is_unloaded) {
     return;
@@ -530,13 +533,24 @@ void tool_unload(bool destruct) {
 
     if (destruct == false) hip_api_trace_buffer.Flush();
 
-    fclose(hip_api_file_handle);
-    fclose(hcc_activity_file_handle);
+    if (hip_api_file_handle != stdout) fclose(hip_api_file_handle);
+    if (hcc_activity_file_handle != stdout) fclose(hcc_activity_file_handle);
   }
+//  printf("TOOL tool_unload end\n"); fflush(stdout);
 }
 
 // HSA-runtime on-unload method
-extern "C" PUBLIC_API void OnUnload() { tool_unload(false); }
+extern "C" PUBLIC_API void OnUnload() {
+//  printf("TOOL OnUnload\n"); fflush(stdout);
+  tool_unload(false);
+//  printf("TOOL OnUnload end\n"); fflush(stdout);
+}
 
-extern "C" CONSTRUCTOR_API void constructor() {}
-extern "C" DESTRUCTOR_API void destructor() { tool_unload(true); }
+extern "C" CONSTRUCTOR_API void constructor() {
+//  printf("TOOL constructor ...end\n"); fflush(stdout);
+}
+extern "C" DESTRUCTOR_API void destructor() {
+//  printf("TOOL destructor\n"); fflush(stdout);
+  tool_unload(true);
+//  printf("TOOL destructor end\n"); fflush(stdout);
+}