contrctor/destructor changes, OnLoad/OnUnload functions, removed hsa_init()/hsa_shutdown()

test/ctrl/test.cpp

@@ -13,6 +13,6 @@ int main(int argc, char** argv) {
   const int diter = (diter_s != NULL) ? atol(diter_s) : 1;
   TestHsa::HsaInstantiate();
   for (int i = 0; i < kiter; ++i) RunKernel<SimpleConvolution, TestAql>(argc, argv, diter);
-  //TestHsa::HsaShutdown();
+  TestHsa::HsaShutdown();
   return 0;
 }

test/ctrl/test_hsa.cpp

@@ -69,7 +69,10 @@ HsaRsrcFactory* TestHsa::HsaInstantiate(const uint32_t agent_ind) {
 }
 
 void TestHsa::HsaShutdown() {
-  if (hsa_queue_ != NULL) hsa_queue_destroy(hsa_queue_);
+  if (hsa_queue_ != NULL) {
+    hsa_queue_destroy(hsa_queue_);
+    hsa_queue_ = NULL;
+  }
   if (hsa_rsrc_) hsa_rsrc_->Destroy();
 }
 

test/ctrl/tool.cpp

@@ -338,7 +338,7 @@ static hsa_status_t info_callback(const rocprofiler_info_data_t info, void * arg
 }
 
 // Tool constructor
-CONSTRUCTOR_API void constructor()
+extern "C" PUBLIC_API void OnLoadTool()
 {
   std::map<std::string, hsa_ven_amd_aqlprofile_parameter_name_t> parameters_dict;
   parameters_dict["HSA_VEN_AMD_AQLPROFILE_PARAMETER_NAME_COMPUTE_UNIT_TARGET"] =
@@ -487,23 +487,21 @@ CONSTRUCTOR_API void constructor()
 }
 
 // Tool destructor
-DESTRUCTOR_API void destructor() {
-  const bool result_file_opened = (result_prefix != NULL) && (result_file_handle != NULL);
+extern "C" PUBLIC_API void OnUnloadTool() {
+  // Unregister dispatch callback
+  rocprofiler_remove_dispatch_callback();
 
+  // Dump stored profiling output data
+  const bool result_file_opened = (result_prefix != NULL) && (result_file_handle != NULL);
   printf("\nROCPRofiler: %u contexts collected", context_count);
   if (result_file_opened) printf(", output directory %s", result_prefix);
   printf("\n");
-
-  // Dump stored profiling output data
   dump_context_array();
+  if (result_file_opened) fclose(result_file_handle);
 
-  // Unregister dispatch callback and free callback data
-  rocprofiler_remove_dispatch_callback();
+  // Cleanup
   if (dispatch_data != NULL) {
     delete[] dispatch_data->features;
     delete dispatch_data;
   }
-
-  // Close output file
-  if (result_file_opened) fclose(result_file_handle);
 }