diff --git a/openmp/libomptarget/plugins/amdgpu/impl/atmi.cpp b/openmp/libomptarget/plugins/amdgpu/impl/atmi.cpp
--- a/openmp/libomptarget/plugins/amdgpu/impl/atmi.cpp
+++ b/openmp/libomptarget/plugins/amdgpu/impl/atmi.cpp
@@ -10,13 +10,6 @@
 #include <hsa_ext_amd.h>
 #include <memory>
 
-/*
- * Initialize/Finalize
- */
-atmi_status_t atmi_init() { return core::Runtime::Initialize(); }
-
-atmi_status_t atmi_finalize() { return core::Runtime::Finalize(); }
-
 /*
  * Machine Info
  */
diff --git a/openmp/libomptarget/plugins/amdgpu/impl/atmi_runtime.h b/openmp/libomptarget/plugins/amdgpu/impl/atmi_runtime.h
--- a/openmp/libomptarget/plugins/amdgpu/impl/atmi_runtime.h
+++ b/openmp/libomptarget/plugins/amdgpu/impl/atmi_runtime.h
@@ -18,38 +18,6 @@
 extern "C" {
 #endif
 
-/** \defgroup context_functions ATMI Context Setup and Finalize
- *  @{
- */
-/**
- * @brief Initialize the ATMI runtime environment.
- *
- * @detal All ATMI runtime functions will fail if this function is not called
- * at least once. The user may initialize difference device types at different
- * regions in the program in order for optimization purposes.
- *
- * @retval ::ATMI_STATUS_SUCCESS The function has executed successfully.
- *
- * @retval ::ATMI_STATUS_ERROR The function encountered errors.
- *
- * @retval ::ATMI_STATUS_UNKNOWN The function encountered errors.
- */
-atmi_status_t atmi_init();
-
-/**
- * @brief Finalize the ATMI runtime environment.
- *
- * @detail ATMI runtime functions will fail if called after finalize.
- *
- * @retval ::ATMI_STATUS_SUCCESS The function has executed successfully.
- *
- * @retval ::ATMI_STATUS_ERROR The function encountered errors.
- *
- * @retval ::ATMI_STATUS_UNKNOWN The function encountered errors.
- */
-atmi_status_t atmi_finalize();
-/** @} */
-
 /** \defgroup module_functions ATMI Module
  * @{
  */
diff --git a/openmp/libomptarget/plugins/amdgpu/impl/rt.h b/openmp/libomptarget/plugins/amdgpu/impl/rt.h
--- a/openmp/libomptarget/plugins/amdgpu/impl/rt.h
+++ b/openmp/libomptarget/plugins/amdgpu/impl/rt.h
@@ -49,9 +49,6 @@
     return instance;
   }
 
-  // init/finalize
-  static atmi_status_t Initialize();
-  static atmi_status_t Finalize();
   // machine info
   static atmi_machine_t *GetMachineInfo();
   // modules
diff --git a/openmp/libomptarget/plugins/amdgpu/impl/system.cpp b/openmp/libomptarget/plugins/amdgpu/impl/system.cpp
--- a/openmp/libomptarget/plugins/amdgpu/impl/system.cpp
+++ b/openmp/libomptarget/plugins/amdgpu/impl/system.cpp
@@ -173,29 +173,6 @@
   return hsa_amd_agents_allow_access(agents.size(), &agents[0], NULL, ptr);
 }
 
-atmi_status_t Runtime::Initialize() {
-  atmi_status_t rc = atl_init_gpu_context();
-  if (rc != ATMI_STATUS_SUCCESS) {
-    printf("[%s:%d] %s failed: %s\n", __FILE__, __LINE__, "GPU context init",
-           get_atmi_error_string(atl_init_gpu_context()));
-    return rc;
-  }
-
-  return ATMI_STATUS_SUCCESS;
-}
-
-atmi_status_t Runtime::Finalize() {
-  atmi_status_t rc = ATMI_STATUS_SUCCESS;
-  hsa_status_t err = hsa_shut_down();
-  if (err != HSA_STATUS_SUCCESS) {
-    printf("[%s:%d] %s failed: %s\n", __FILE__, __LINE__, "Shutting down HSA",
-           get_error_string(err));
-    rc = ATMI_STATUS_ERROR;
-  }
-
-  return rc;
-}
-
 static void atmi_init_context_structs() {
   atlc.struct_initialized = true; /* This only gets called one time */
   atlc.g_hsa_initialized = false;
diff --git a/openmp/libomptarget/plugins/amdgpu/src/rtl.cpp b/openmp/libomptarget/plugins/amdgpu/src/rtl.cpp
--- a/openmp/libomptarget/plugins/amdgpu/src/rtl.cpp
+++ b/openmp/libomptarget/plugins/amdgpu/src/rtl.cpp
@@ -464,11 +464,12 @@
       print_kernel_trace = 0;
 
     DP("Start initializing HSA-ATMI\n");
-    atmi_status_t err = atmi_init();
+    atmi_status_t err = core::atl_init_gpu_context();
     if (err != ATMI_STATUS_SUCCESS) {
       DP("Error when initializing HSA-ATMI\n");
       return;
     }
+
     // Init hostcall soon after initializing ATMI
     hostrpc_init();
 
@@ -579,15 +580,20 @@
     // Terminate hostrpc before finalizing ATMI
     hostrpc_terminate();
 
+    hsa_status_t Err;
     for (uint32_t I = 0; I < HSAExecutables.size(); I++) {
-      hsa_status_t Err = hsa_executable_destroy(HSAExecutables[I]);
+      Err = hsa_executable_destroy(HSAExecutables[I]);
       if (Err != HSA_STATUS_SUCCESS) {
         DP("[%s:%d] %s failed: %s\n", __FILE__, __LINE__,
            "Destroying executable", get_error_string(Err));
       }
     }
 
-    atmi_finalize();
+    Err = hsa_shut_down();
+    if (Err != HSA_STATUS_SUCCESS) {
+      printf("[%s:%d] %s failed: %s\n", __FILE__, __LINE__, "Shutting down HSA",
+             get_error_string(Err));
+    }
   }
 };