1 // omp_target_disassociate_ptr should always fail if the hold reference count is
2 // non-zero, regardless of the dynamic reference count. When the latter is
3 // finite, the implementation happens to choose to report the hold diagnostic.
4
5 // RUN: %libomptarget-compile-generic -fopenmp-extensions
6 // RUN: %not %libomptarget-run-generic 0 2>&1 | %fcheck-generic
7 // RUN: %not %libomptarget-run-generic 1 2>&1 | %fcheck-generic
8 // RUN: %not %libomptarget-run-generic inf 2>&1 | %fcheck-generic
9
10 // RUN: %libomptarget-compile-generic -fopenmp-extensions -DHOLD_MORE
11 // RUN: %not %libomptarget-run-generic 0 2>&1 | %fcheck-generic
12 // RUN: %not %libomptarget-run-generic 1 2>&1 | %fcheck-generic
13 // RUN: %not %libomptarget-run-generic inf 2>&1 | %fcheck-generic
14
15 #include <limits.h>
16 #include <omp.h>
17 #include <stdio.h>
18 #include <string.h>
19
main(int argc,char * argv[])20 int main(int argc, char *argv[]) {
21 // Parse command line.
22 int DynRef;
23 if (argc != 2) {
24 fprintf(stderr, "bad arguments\n");
25 return 1;
26 }
27 if (0 == strcmp(argv[1], "inf"))
28 DynRef = INT_MAX;
29 else
30 DynRef = atoi(argv[1]);
31
32 // Allocate and set dynamic reference count as specified.
33 int DevNum = omp_get_default_device();
34 int X;
35 void *XDev = omp_target_alloc(sizeof X, DevNum);
36 if (!XDev) {
37 fprintf(stderr, "omp_target_alloc failed\n");
38 return 1;
39 }
40 if (DynRef == INT_MAX) {
41 if (omp_target_associate_ptr(&X, &XDev, sizeof X, 0, DevNum)) {
42 fprintf(stderr, "omp_target_associate_ptr failed\n");
43 return 1;
44 }
45 } else {
46 for (int I = 0; I < DynRef; ++I) {
47 #pragma omp target enter data map(alloc : X)
48 }
49 }
50
51 // Disassociate while hold reference count > 0.
52 int Status = 0;
53 #pragma omp target data map(ompx_hold, alloc : X)
54 #if HOLD_MORE
55 #pragma omp target data map(ompx_hold, alloc : X)
56 #pragma omp target data map(ompx_hold, alloc : X)
57 #endif
58 {
59 // CHECK: omptarget error: Trying to disassociate a pointer with a
60 // CHECK-SAME: non-zero hold reference count
61 // CHECK-NEXT: omp_target_disassociate_ptr failed
62 if (omp_target_disassociate_ptr(&X, DevNum)) {
63 fprintf(stderr, "omp_target_disassociate_ptr failed\n");
64 Status = 1;
65 }
66 }
67 return Status;
68 }
69