aboutsummaryrefslogtreecommitdiff
diff options
context:
space:
mode:
authorThomas Schwinge <tschwinge@baylibre.com>2025-04-23 17:35:29 +0200
committerSandra Loosemore <sloosemore@baylibre.com>2025-05-15 20:25:54 +0000
commitc4362bb0317a21e6b8ab4003d78e39bbfd2a5e7f (patch)
treed70d7e9aed692223349c016f79543ab13d78d8e8
parenta9b364c4ab1e1617196d38c087b71507710882e1 (diff)
downloadgcc-c4362bb0317a21e6b8ab4003d78e39bbfd2a5e7f.zip
gcc-c4362bb0317a21e6b8ab4003d78e39bbfd2a5e7f.tar.gz
gcc-c4362bb0317a21e6b8ab4003d78e39bbfd2a5e7f.tar.bz2
Add 'libgomp.c-c++-common/target-cdtor-1.c'
libgomp/ * testsuite/libgomp.c-c++-common/target-cdtor-1.c: New. (cherry picked from commit 40ce48e87c1e7344c622c8eb6bed53f1311f5a0a)
-rw-r--r--libgomp/testsuite/libgomp.c-c++-common/target-cdtor-1.c89
1 files changed, 89 insertions, 0 deletions
diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-cdtor-1.c b/libgomp/testsuite/libgomp.c-c++-common/target-cdtor-1.c
new file mode 100644
index 0000000..e6099cf
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/target-cdtor-1.c
@@ -0,0 +1,89 @@
+/* Offloaded 'constructor' and 'destructor' functions. */
+
+#include <omp.h>
+
+#pragma omp declare target
+
+static void
+__attribute__((constructor))
+initHD1()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+static void
+__attribute__((constructor))
+initHD2()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+static void
+__attribute__((destructor))
+finiHD1()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+static void
+__attribute__((destructor))
+finiHD2()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+#pragma omp end declare target
+
+static void
+__attribute__((constructor))
+initH1()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+static void
+__attribute__((destructor))
+finiH2()
+{
+ __builtin_printf("%s, %d\n", __FUNCTION__, omp_is_initial_device());
+}
+
+int main()
+{
+ int c = 0;
+
+ __builtin_printf("%s:%d, %d\n", __FUNCTION__, ++c, omp_is_initial_device());
+
+#pragma omp target map(c)
+ {
+ __builtin_printf("%s:%d, %d\n", __FUNCTION__, ++c, omp_is_initial_device());
+ }
+
+#pragma omp target map(c)
+ {
+ __builtin_printf("%s:%d, %d\n", __FUNCTION__, ++c, omp_is_initial_device());
+ }
+
+ __builtin_printf("%s:%d, %d\n", __FUNCTION__, ++c, omp_is_initial_device());
+
+ return 0;
+}
+
+/* The order is undefined, in which same-priority 'constructor' functions, and 'destructor' functions are run.
+ { dg-output {init[^,]+, 1[\r\n]+} }
+ { dg-output {init[^,]+, 1[\r\n]+} }
+ { dg-output {init[^,]+, 1[\r\n]+} }
+ { dg-output {main:1, 1[\r\n]+} }
+ { dg-output {initHD[^,]+, 0[\r\n]+} { target offload_device } }
+ { dg-output {initHD[^,]+, 0[\r\n]+} { target offload_device } }
+ { dg-output {main:2, 1[\r\n]+} { target { ! offload_device } } }
+ { dg-output {main:2, 0[\r\n]+} { target offload_device } }
+ { dg-output {main:3, 1[\r\n]+} { target { ! offload_device } } }
+ { dg-output {main:3, 0[\r\n]+} { target offload_device } }
+ { dg-output {main:4, 1[\r\n]+} }
+ { dg-output {finiHD[^,]+, 0[\r\n]+} { target offload_device } }
+ { dg-output {finiHD[^,]+, 0[\r\n]+} { target offload_device } }
+ { dg-output {fini[^,]+, 1[\r\n]+} }
+ { dg-output {fini[^,]+, 1[\r\n]+} }
+ { dg-output {fini[^,]+, 1[\r\n]+} }
+*/