diff --git a/openmp/libomptarget/CMakeLists.txt b/openmp/libomptarget/CMakeLists.txt
--- a/openmp/libomptarget/CMakeLists.txt
+++ b/openmp/libomptarget/CMakeLists.txt
@@ -37,6 +37,7 @@
 
 # This is a list of all the targets that are supported/tested right now.
 set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} aarch64-unknown-linux-gnu")
+set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} amdgcn-amd-amdhsa")
 set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} powerpc64le-ibm-linux-gnu")
 set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} powerpc64-ibm-linux-gnu")
 set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} x86_64-pc-linux-gnu")
diff --git a/openmp/libomptarget/deviceRTLs/CMakeLists.txt b/openmp/libomptarget/deviceRTLs/CMakeLists.txt
--- a/openmp/libomptarget/deviceRTLs/CMakeLists.txt
+++ b/openmp/libomptarget/deviceRTLs/CMakeLists.txt
@@ -9,5 +9,5 @@
 # Build a device RTL for each available machine.
 #
 ##===----------------------------------------------------------------------===##
-
+add_subdirectory(amdgcn)
 add_subdirectory(nvptx)
diff --git a/openmp/libomptarget/test/env/base_ptr_ref_count.c b/openmp/libomptarget/test/env/base_ptr_ref_count.c
--- a/openmp/libomptarget/test/env/base_ptr_ref_count.c
+++ b/openmp/libomptarget/test/env/base_ptr_ref_count.c
@@ -3,6 +3,8 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-powerpc64le-ibm-linux-gnu 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa
+
 // REQUIRES: libomptarget-debug
 
 #include <stdlib.h>
diff --git a/openmp/libomptarget/test/env/omp_target_debug.c b/openmp/libomptarget/test/env/omp_target_debug.c
--- a/openmp/libomptarget/test/env/omp_target_debug.c
+++ b/openmp/libomptarget/test/env/omp_target_debug.c
@@ -8,6 +8,8 @@
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env LIBOMPTARGET_DEBUG=0 %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu -allow-empty -check-prefix=NDEBUG
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=0 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=NDEBUG
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa -allow-empty -check-prefix=DEBUG
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=0 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa -allow-empty -check-prefix=NDEBUG
 // REQUIRES: libomptarget-debug
 
 int main(void) {
diff --git a/openmp/libomptarget/test/mapping/data_absent_at_exit.c b/openmp/libomptarget/test/mapping/data_absent_at_exit.c
--- a/openmp/libomptarget/test/mapping/data_absent_at_exit.c
+++ b/openmp/libomptarget/test/mapping/data_absent_at_exit.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <stdio.h>
 
diff --git a/openmp/libomptarget/test/mapping/declare_mapper_api.cpp b/openmp/libomptarget/test/mapping/declare_mapper_api.cpp
--- a/openmp/libomptarget/test/mapping/declare_mapper_api.cpp
+++ b/openmp/libomptarget/test/mapping/declare_mapper_api.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-run-and-check-amdgcn-amd-amdhsa
 
 #include <cstdio>
 #include <cstdlib>
diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target.cpp
--- a/openmp/libomptarget/test/mapping/declare_mapper_target.cpp
+++ b/openmp/libomptarget/test/mapping/declare_mapper_target.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <cstdio>
 #include <cstdlib>
diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp
--- a/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp
+++ b/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <cstdio>
 #include <cstdlib>
diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp
--- a/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp
+++ b/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <cstdio>
 #include <cstdlib>
diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp
--- a/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp
+++ b/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <cstdio>
 #include <cstdlib>
diff --git a/openmp/libomptarget/test/mapping/lambda_mapping.cpp b/openmp/libomptarget/test/mapping/lambda_mapping.cpp
--- a/openmp/libomptarget/test/mapping/lambda_mapping.cpp
+++ b/openmp/libomptarget/test/mapping/lambda_mapping.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-run-and-check-amdgcn-amd-amdhsa
 
 #include <iostream>
 
diff --git a/openmp/libomptarget/test/mapping/pr38704.c b/openmp/libomptarget/test/mapping/pr38704.c
--- a/openmp/libomptarget/test/mapping/pr38704.c
+++ b/openmp/libomptarget/test/mapping/pr38704.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 // Clang 6.0 doesn't use the new map interface, undefined behavior when
 // the compiler emits "old" interface code for structures.
diff --git a/openmp/libomptarget/test/mapping/private_mapping.c b/openmp/libomptarget/test/mapping/private_mapping.c
--- a/openmp/libomptarget/test/mapping/private_mapping.c
+++ b/openmp/libomptarget/test/mapping/private_mapping.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <assert.h>
 #include <stdio.h>
diff --git a/openmp/libomptarget/test/offloading/assert.cpp b/openmp/libomptarget/test/offloading/assert.cpp
--- a/openmp/libomptarget/test/offloading/assert.cpp
+++ b/openmp/libomptarget/test/offloading/assert.cpp
@@ -1,4 +1,5 @@
 // RUN: %libomptarget-compilexx-nvptx64-nvidia-cuda && %libomptarget-run-fail-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-amdgcn-amd-amdhsa && %libomptarget-run-fail-amdgcn-amd-amdhsa
 
 int main(int argc, char *argv[]) {
 #pragma omp target
diff --git a/openmp/libomptarget/test/offloading/bug47654.cpp b/openmp/libomptarget/test/offloading/bug47654.cpp
--- a/openmp/libomptarget/test/offloading/bug47654.cpp
+++ b/openmp/libomptarget/test/offloading/bug47654.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-run-and-check-amdgcn-amd-amdhsa
 
 #include <cassert>
 #include <iostream>
diff --git a/openmp/libomptarget/test/offloading/d2d_memcpy.c b/openmp/libomptarget/test/offloading/d2d_memcpy.c
--- a/openmp/libomptarget/test/offloading/d2d_memcpy.c
+++ b/openmp/libomptarget/test/offloading/d2d_memcpy.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu && env OMP_MAX_ACTIVE_LEVELS=2 %libomptarget-run-powerpc64le-ibm-linux-gnu | %fcheck-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env OMP_MAX_ACTIVE_LEVELS=2 %libomptarget-run-x86_64-pc-linux-gnu | %fcheck-x86_64-pc-linux-gnu -allow-empty
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env OMP_MAX_ACTIVE_LEVELS=2 %libomptarget-run-nvptx64-nvidia-cuda | %fcheck-nvptx64-nvidia-cuda -allow-empty
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env OMP_MAX_ACTIVE_LEVELS=2 %libomptarget-run-amdgcn-amd-amdhsa | %fcheck-amdgcn-amd-amdhsa -allow-empty
 
 #include <assert.h>
 #include <omp.h>
diff --git a/openmp/libomptarget/test/offloading/dynamic_module.c b/openmp/libomptarget/test/offloading/dynamic_module.c
--- a/openmp/libomptarget/test/offloading/dynamic_module.c
+++ b/openmp/libomptarget/test/offloading/dynamic_module.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu -DSHARED -fPIC -shared -o %t.so && %libomptarget-compile-powerpc64le-ibm-linux-gnu %t.so && %libomptarget-run-powerpc64le-ibm-linux-gnu 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu -DSHARED -fPIC -shared -o %t.so && %libomptarget-compile-x86_64-pc-linux-gnu %t.so && %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda -DSHARED -fPIC -shared -o %t.so && %libomptarget-compile-nvptx64-nvidia-cuda %t.so && %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -DSHARED -fPIC -shared -o %t.so && %libomptarget-compile-amdgcn-amd-amdhsa %t.so && %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa
 
 #ifdef SHARED
 void foo() {}
diff --git a/openmp/libomptarget/test/offloading/dynamic_module_load.c b/openmp/libomptarget/test/offloading/dynamic_module_load.c
--- a/openmp/libomptarget/test/offloading/dynamic_module_load.c
+++ b/openmp/libomptarget/test/offloading/dynamic_module_load.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t-powerpc64le-ibm-linux-gnu -ldl && %libomptarget-run-powerpc64le-ibm-linux-gnu %t.so 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t-x86_64-pc-linux-gnu -ldl && %libomptarget-run-x86_64-pc-linux-gnu %t.so 2>&1 | %fcheck-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t-nvptx64-nvidia-cuda -ldl && %libomptarget-run-nvptx64-nvidia-cuda %t.so 2>&1 | %fcheck-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa -DSHARED -fPIC -shared -o %t.so && %clang %flags %s -o %t-amdgcn-amd-amdhsa -ldl && %libomptarget-run-amdgcn-amd-amdhsa %t.so 2>&1 | %fcheck-amdgcn-amd-amdhsa
 
 #ifdef SHARED
 #include <stdio.h>
diff --git a/openmp/libomptarget/test/offloading/lone_target_exit_data.c b/openmp/libomptarget/test/offloading/lone_target_exit_data.c
--- a/openmp/libomptarget/test/offloading/lone_target_exit_data.c
+++ b/openmp/libomptarget/test/offloading/lone_target_exit_data.c
@@ -6,6 +6,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <stdio.h>
 
diff --git a/openmp/libomptarget/test/offloading/looptripcnt.c b/openmp/libomptarget/test/offloading/looptripcnt.c
--- a/openmp/libomptarget/test/offloading/looptripcnt.c
+++ b/openmp/libomptarget/test/offloading/looptripcnt.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-powerpc64le-ibm-linux-gnu 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=DEBUG
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa -allow-empty -check-prefix=DEBUG
 // REQUIRES: libomptarget-debug
 
 /*
diff --git a/openmp/libomptarget/test/offloading/non_contiguous_update.cpp b/openmp/libomptarget/test/offloading/non_contiguous_update.cpp
--- a/openmp/libomptarget/test/offloading/non_contiguous_update.cpp
+++ b/openmp/libomptarget/test/offloading/non_contiguous_update.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-powerpc64le-ibm-linux-gnu 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=DEBUG
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa -allow-empty -check-prefix=DEBUG
 // REQUIRES: libomptarget-debug
 
 #include <cstdio>
diff --git a/openmp/libomptarget/test/offloading/offloading_success.c b/openmp/libomptarget/test/offloading/offloading_success.c
--- a/openmp/libomptarget/test/offloading/offloading_success.c
+++ b/openmp/libomptarget/test/offloading/offloading_success.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compile-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compile-run-and-check-amdgcn-amd-amdhsa
 
 #include <stdio.h>
 #include <omp.h>
diff --git a/openmp/libomptarget/test/offloading/offloading_success.cpp b/openmp/libomptarget/test/offloading/offloading_success.cpp
--- a/openmp/libomptarget/test/offloading/offloading_success.cpp
+++ b/openmp/libomptarget/test/offloading/offloading_success.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-run-and-check-amdgcn-amd-amdhsa
 
 #include <stdio.h>
 #include <omp.h>
diff --git a/openmp/libomptarget/test/offloading/requires.c b/openmp/libomptarget/test/offloading/requires.c
--- a/openmp/libomptarget/test/offloading/requires.c
+++ b/openmp/libomptarget/test/offloading/requires.c
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-powerpc64le-ibm-linux-gnu 2>&1 | %fcheck-powerpc64le-ibm-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-x86_64-pc-linux-gnu && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-x86_64-pc-linux-gnu 2>&1 | %fcheck-x86_64-pc-linux-gnu -allow-empty -check-prefix=DEBUG
 // RUN: %libomptarget-compile-nvptx64-nvidia-cuda && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-nvptx64-nvidia-cuda 2>&1 | %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=DEBUG
+// RUN: %libomptarget-compile-amdgcn-amd-amdhsa && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-amdgcn-amd-amdhsa 2>&1 | %fcheck-amdgcn-amd-amdhsa -allow-empty -check-prefix=DEBUG
 // REQUIRES: libomptarget-debug
 
 /*
diff --git a/openmp/libomptarget/test/offloading/target_depend_nowait.cpp b/openmp/libomptarget/test/offloading/target_depend_nowait.cpp
--- a/openmp/libomptarget/test/offloading/target_depend_nowait.cpp
+++ b/openmp/libomptarget/test/offloading/target_depend_nowait.cpp
@@ -3,6 +3,7 @@
 // RUN: %libomptarget-compilexx-run-and-check-powerpc64le-ibm-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-x86_64-pc-linux-gnu
 // RUN: %libomptarget-compilexx-run-and-check-nvptx64-nvidia-cuda
+// RUN: %libomptarget-compilexx-run-and-check-amdgcn-amd-amdhsa
 
 #include <omp.h>
 #include <stdio.h>