diff options
| author | Joseph Huber <jhuber6@vols.utk.edu> | 2022-10-19 16:26:35 +0300 |
|---|---|---|
| committer | Joseph Huber <jhuber6@vols.utk.edu> | 2022-10-19 16:57:27 +0300 |
| commit | 586fc5999bb7466451ebf4ce9183717679144304 (patch) | |
| tree | 680b212497d73dee2481dfb77eb99de658ee913a /openmp | |
| parent | ad6698562c3d43d0311922e7f6aebb8998021530 (diff) | |
[Libomptarget][NFC] clang-format the libomptarget OpenMP tests
Summary:
Recent changes to clang-format improved the handling of OpenMP pragmas.
Clean up the existing libomptarget tests.
Diffstat (limited to 'openmp')
80 files changed, 470 insertions, 515 deletions
diff --git a/openmp/libomptarget/test/api/omp_get_device_num.c b/openmp/libomptarget/test/api/omp_get_device_num.c index a0a897a20479..31e13b0050d4 100644 --- a/openmp/libomptarget/test/api/omp_get_device_num.c +++ b/openmp/libomptarget/test/api/omp_get_device_num.c @@ -1,24 +1,22 @@ // RUN: %libomptarget-compile-run-and-check-generic -#include <stdio.h> #include <omp.h> +#include <stdio.h> -int test_omp_get_device_num() -{ +int test_omp_get_device_num() { /* checks that omp_get_device_num() == omp_get_num_devices() in the host */ int device_num = omp_get_device_num(); printf("device_num = %d\n", device_num); - #pragma omp target +#pragma omp target {} return (device_num == omp_get_num_devices()); } -int main() -{ +int main() { int i; - int failed=0; + int failed = 0; if (!test_omp_get_device_num()) { failed++; diff --git a/openmp/libomptarget/test/api/omp_get_num_devices.c b/openmp/libomptarget/test/api/omp_get_num_devices.c index c2d455fec2af..e8fbdcd1a1de 100644 --- a/openmp/libomptarget/test/api/omp_get_num_devices.c +++ b/openmp/libomptarget/test/api/omp_get_num_devices.c @@ -1,24 +1,22 @@ // RUN: %libomptarget-compile-run-and-check-generic -#include <stdio.h> #include <omp.h> +#include <stdio.h> -int test_omp_get_num_devices() -{ +int test_omp_get_num_devices() { /* checks that omp_get_num_devices() > 0 */ int num_devices = omp_get_num_devices(); printf("num_devices = %d\n", num_devices); - #pragma omp target +#pragma omp target {} return (num_devices > 0); } -int main() -{ +int main() { int i; - int failed=0; + int failed = 0; if (!test_omp_get_num_devices()) { failed++; diff --git a/openmp/libomptarget/test/api/omp_host_pinned_memory.c b/openmp/libomptarget/test/api/omp_host_pinned_memory.c index 8531dad19010..7a6a00d489d5 100644 --- a/openmp/libomptarget/test/api/omp_host_pinned_memory.c +++ b/openmp/libomptarget/test/api/omp_host_pinned_memory.c @@ -17,8 +17,8 @@ int main() { for (int i = 0; i < N; ++i) hst_ptr[i] = 2; -#pragma omp target teams distribute parallel for device(device) \ - map(tofrom:hst_ptr[0 : N]) +#pragma omp target teams distribute parallel for device(device) \ + map(tofrom : hst_ptr[0 : N]) for (int i = 0; i < N; ++i) hst_ptr[i] -= 1; @@ -29,5 +29,5 @@ int main() { llvm_omp_target_free_host(hst_ptr, device); // CHECK: PASS if (sum == N) - printf ("PASS\n"); + printf("PASS\n"); } diff --git a/openmp/libomptarget/test/api/omp_host_pinned_memory_alloc.c b/openmp/libomptarget/test/api/omp_host_pinned_memory_alloc.c index c83bb23bd807..0b8118fa219a 100644 --- a/openmp/libomptarget/test/api/omp_host_pinned_memory_alloc.c +++ b/openmp/libomptarget/test/api/omp_host_pinned_memory_alloc.c @@ -11,7 +11,7 @@ int main() { for (int i = 0; i < N; ++i) hst_ptr[i] = 2; -#pragma omp target teams distribute parallel for map(tofrom : hst_ptr [0:N]) +#pragma omp target teams distribute parallel for map(tofrom : hst_ptr[0 : N]) for (int i = 0; i < N; ++i) hst_ptr[i] -= 1; diff --git a/openmp/libomptarget/test/env/base_ptr_ref_count.c b/openmp/libomptarget/test/env/base_ptr_ref_count.c index bd2166e8b384..9a1556871280 100644 --- a/openmp/libomptarget/test/env/base_ptr_ref_count.c +++ b/openmp/libomptarget/test/env/base_ptr_ref_count.c @@ -1,25 +1,23 @@ // RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic // REQUIRES: libomptarget-debug -#include <stdlib.h> #include <stdio.h> +#include <stdlib.h> int *allocate(size_t n) { int *ptr = malloc(sizeof(int) * n); -#pragma omp target enter data map(to : ptr[:n]) +#pragma omp target enter data map(to : ptr[ : n]) return ptr; } void deallocate(int *ptr, size_t n) { -#pragma omp target exit data map(delete : ptr[:n]) +#pragma omp target exit data map(delete : ptr[ : n]) free(ptr); } #pragma omp declare target int *cnt; -void foo() { - ++(*cnt); -} +void foo() { ++(*cnt); } #pragma omp end declare target int main(void) { @@ -27,21 +25,20 @@ int main(void) { int *V = allocate(10); deallocate(A, 10); deallocate(V, 10); -// CHECK-NOT: RefCount=2 + // CHECK-NOT: RefCount=2 cnt = malloc(sizeof(int)); *cnt = 0; -#pragma omp target map(cnt[:1]) +#pragma omp target map(cnt[ : 1]) foo(); printf("Cnt = %d.\n", *cnt); -// CHECK: Cnt = 1. + // CHECK: Cnt = 1. *cnt = 0; -#pragma omp target data map(cnt[:1]) +#pragma omp target data map(cnt[ : 1]) #pragma omp target foo(); printf("Cnt = %d.\n", *cnt); -// CHECK: Cnt = 1. + // CHECK: Cnt = 1. free(cnt); return 0; } - diff --git a/openmp/libomptarget/test/env/omp_target_debug.c b/openmp/libomptarget/test/env/omp_target_debug.c index 8793d38df718..76d182d2481c 100644 --- a/openmp/libomptarget/test/env/omp_target_debug.c +++ b/openmp/libomptarget/test/env/omp_target_debug.c @@ -11,4 +11,3 @@ int main(void) { // DEBUG: Libomptarget // NDEBUG-NOT: Libomptarget // NDEBUG-NOT: Target - diff --git a/openmp/libomptarget/test/mapping/alloc_fail.c b/openmp/libomptarget/test/mapping/alloc_fail.c index 7e22c21627b3..d764087723ba 100644 --- a/openmp/libomptarget/test/mapping/alloc_fail.c +++ b/openmp/libomptarget/test/mapping/alloc_fail.c @@ -8,8 +8,8 @@ int main() { int arr[4] = {0, 1, 2, 3}; -#pragma omp target data map(alloc: arr[0:2]) -#pragma omp target data map(alloc: arr[1:2]) +#pragma omp target data map(alloc : arr[0 : 2]) +#pragma omp target data map(alloc : arr[1 : 2]) ; return 0; } diff --git a/openmp/libomptarget/test/mapping/array_section_implicit_capture.c b/openmp/libomptarget/test/mapping/array_section_implicit_capture.c index 04f0f1167ae0..f049a84da2bf 100644 --- a/openmp/libomptarget/test/mapping/array_section_implicit_capture.c +++ b/openmp/libomptarget/test/mapping/array_section_implicit_capture.c @@ -23,8 +23,8 @@ int main() { B[i] = 2 * i; } -#pragma omp target enter data map(to : A [FROM:LENGTH], B [FROM:LENGTH]) -#pragma omp target enter data map(alloc : C [FROM:LENGTH]) +#pragma omp target enter data map(to : A[FROM : LENGTH], B[FROM : LENGTH]) +#pragma omp target enter data map(alloc : C[FROM : LENGTH]) // A, B and C have been mapped starting at index FROM, but inside the kernel // they are captured implicitly so the library must look them up using their @@ -36,8 +36,8 @@ int main() { } } -#pragma omp target exit data map(from : C [FROM:LENGTH]) -#pragma omp target exit data map(delete : A [FROM:LENGTH], B [FROM:LENGTH]) +#pragma omp target exit data map(from : C[FROM : LENGTH]) +#pragma omp target exit data map(delete : A[FROM : LENGTH], B[FROM : LENGTH]) int errors = 0; for (int i = FROM; i < FROM + LENGTH; i++) diff --git a/openmp/libomptarget/test/mapping/array_section_use_device_ptr.c b/openmp/libomptarget/test/mapping/array_section_use_device_ptr.c index ca63eb9cd98e..17a81b91882a 100644 --- a/openmp/libomptarget/test/mapping/array_section_use_device_ptr.c +++ b/openmp/libomptarget/test/mapping/array_section_use_device_ptr.c @@ -12,7 +12,7 @@ int main() { float *A = (float *)malloc(N * sizeof(float)); -#pragma omp target enter data map(to : A [FROM:LENGTH]) +#pragma omp target enter data map(to : A[FROM : LENGTH]) // A, has been mapped starting at index FROM, but inside the use_device_ptr // clause it is captured by base so the library must look it up using the @@ -21,7 +21,7 @@ int main() { float *A_dev = NULL; #pragma omp target data use_device_ptr(A) { A_dev = A; } -#pragma omp target exit data map(delete : A [FROM:LENGTH]) +#pragma omp target exit data map(delete : A[FROM : LENGTH]) // CHECK: Success if (A_dev == NULL || A_dev == A) diff --git a/openmp/libomptarget/test/mapping/data_absent_at_exit.c b/openmp/libomptarget/test/mapping/data_absent_at_exit.c index a4e31cf20458..5baaf933b283 100644 --- a/openmp/libomptarget/test/mapping/data_absent_at_exit.c +++ b/openmp/libomptarget/test/mapping/data_absent_at_exit.c @@ -11,18 +11,17 @@ int main(void) { int f = 5, r = 6, d = 7, af = 8; - // Check exit from omp target data. - // CHECK: f = 5, af = 8 - #pragma omp target data map(from: f) map(always, from: af) +// Check exit from omp target data. +// CHECK: f = 5, af = 8 +#pragma omp target data map(from : f) map(always, from : af) { - #pragma omp target exit data map(delete: f, af) - } - printf("f = %d, af = %d\n", f, af); +#pragma omp target exit data map(delete : f, af) + } printf("f = %d, af = %d\n", f, af); - // Check omp target exit data. - // CHECK: f = 5, r = 6, d = 7, af = 8 - #pragma omp target exit data map(from: f) map(release: r) map(delete: d) \ - map(always, from: af) +// Check omp target exit data. +// CHECK: f = 5, r = 6, d = 7, af = 8 +#pragma omp target exit data map(from : f) map(release : r) map(delete : d) \ + map(always, from : af) printf("f = %d, r = %d, d = %d, af = %d\n", f, r, d, af); return 0; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_api.cpp b/openmp/libomptarget/test/mapping/declare_mapper_api.cpp index 90c3c032be9e..041c7f64be21 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_api.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_api.cpp @@ -1,9 +1,9 @@ // RUN: %libomptarget-compilexx-run-and-check-generic +#include <cinttypes> #include <cstdio> #include <cstdlib> #include <vector> -#include <cinttypes> // Data structure definitions copied from OpenMP RTL. struct MapComponentInfoTy { @@ -13,7 +13,8 @@ struct MapComponentInfoTy { int64_t Type; void *Name; MapComponentInfoTy() = default; - MapComponentInfoTy(void *Base, void *Begin, int64_t Size, int64_t Type, void *Name) + MapComponentInfoTy(void *Base, void *Begin, int64_t Size, int64_t Type, + void *Name) : Base(Base), Begin(Begin), Size(Size), Type(Type), Name(Name) {} }; @@ -27,7 +28,7 @@ extern "C" { #endif int64_t __tgt_mapper_num_components(void *rt_mapper_handle); void __tgt_push_mapper_component(void *rt_mapper_handle, void *base, - void *begin, int64_t size, int64_t type, + void *begin, int64_t size, int64_t type, void *name); #ifdef __cplusplus } diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers.cpp index 32e552a38c83..f7f92591988b 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers.cpp @@ -11,14 +11,14 @@ typedef struct { int a; double *b; } C1; -#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int a; double *b; C1 c; } C; -#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b [0:2]) +#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b[0 : 2]) typedef struct { int e; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array.cpp index 87d2ef5fc4f3..dddbd332eb74 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array.cpp @@ -13,14 +13,14 @@ typedef struct { int a; double *b; } C1; -#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int a; double *b; C1 c; } C; -#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b [0:2]) +#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b[0 : 2]) typedef struct { int e; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp index e7242aceb1d5..bf510763526f 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_array_subscript.cpp @@ -11,14 +11,14 @@ typedef struct { int a; double *b; } C1; -#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int a; double *b; C1 c; } C; -#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b [0:2]) +#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b[0 : 2]) typedef struct { int e; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp index 968e2dfb2365..d6092213a8f9 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_complex_structure.cpp @@ -47,7 +47,7 @@ public: MyObjectA *arr; int len; }; -#pragma omp declare mapper(MyObjectB obj) map(obj, obj.arr[:obj.len]) +#pragma omp declare mapper(MyObjectB obj) map(obj, obj.arr[ : obj.len]) class MyObjectC { public: @@ -67,7 +67,7 @@ public: MyObjectB *arr; int len; }; -#pragma omp declare mapper(MyObjectC obj) map(obj, obj.arr[:obj.len]) +#pragma omp declare mapper(MyObjectC obj) map(obj, obj.arr[ : obj.len]) int main(void) { MyObjectC *outer = new MyObjectC[N]; @@ -77,7 +77,7 @@ int main(void) { outer[i].show(); printf("Sending data to device...\n"); -#pragma omp target enter data map(to : outer[:N]) +#pragma omp target enter data map(to : outer[ : N]) printf("Calling foo()...\n"); #pragma omp target teams distribute parallel for @@ -87,7 +87,7 @@ int main(void) { printf("foo() complete!\n"); printf("Sending data back to host...\n"); -#pragma omp target exit data map(from : outer[:N]) +#pragma omp target exit data map(from : outer[ : N]) printf("Modified Data Hierarchy:\n"); for (int i = 0; i < N; i++) diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp index 8847919a6235..27ea747e34e2 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_ptr_subscript.cpp @@ -11,14 +11,14 @@ typedef struct { int a; double *b; } C1; -#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int a; double *b; C1 c; } C; -#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b [0:2]) +#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b[0 : 2]) typedef struct { int e; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_var.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_var.cpp index 20a907ea9588..b7f6ef5642e2 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_var.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_default_mappers_var.cpp @@ -11,14 +11,14 @@ typedef struct { int a; double *b; } C1; -#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(C1 s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int a; double *b; C1 c; } C; -#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b [0:2]) +#pragma omp declare mapper(C s) map(to : s.a, s.c) map(from : s.b[0 : 2]) typedef struct { int e; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_nested_mappers.cpp b/openmp/libomptarget/test/mapping/declare_mapper_nested_mappers.cpp index c8956345573a..ce209d5b82a0 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_nested_mappers.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_nested_mappers.cpp @@ -11,7 +11,7 @@ typedef struct { int a; double *b; } C; -#pragma omp declare mapper(id1 : C s) map(to : s.a) map(from : s.b [0:2]) +#pragma omp declare mapper(id1 : C s) map(to : s.a) map(from : s.b[0 : 2]) typedef struct { int e; @@ -19,11 +19,8 @@ typedef struct { int h; short *g; } D; -#pragma omp declare mapper(default \ - : D r) map(from \ - : r.e) map(mapper(id1), tofrom \ - : r.f) map(tofrom \ - : r.g [0:r.h]) +#pragma omp declare mapper(default : D r) map(from : r.e) \ + map(mapper(id1), tofrom : r.f) map(tofrom : r.g[0 : r.h]) int main() { constexpr int N = 10; diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target.cpp index cb999aae2424..4d7237e94657 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_target.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_target.cpp @@ -10,15 +10,15 @@ public: int *a; }; -#pragma omp declare mapper(id: C s) map(s.a[0:NUM]) +#pragma omp declare mapper(id : C s) map(s.a[0 : NUM]) int main() { C c; - c.a = (int*) malloc(sizeof(int)*NUM); + c.a = (int *)malloc(sizeof(int) * NUM); for (int i = 0; i < NUM; i++) { c.a[i] = 1; } - #pragma omp target teams distribute parallel for map(mapper(id),tofrom: c) +#pragma omp target teams distribute parallel for map(mapper(id), tofrom : c) for (int i = 0; i < NUM; i++) { ++c.a[i]; } @@ -30,4 +30,3 @@ int main() { printf("Sum = %d\n", sum); return 0; } - diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp index c5b90ed3253c..7f0984401400 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_target_data.cpp @@ -10,17 +10,17 @@ public: int *a; }; -#pragma omp declare mapper(id: C s) map(s.a[0:NUM]) +#pragma omp declare mapper(id : C s) map(s.a[0 : NUM]) int main() { C c; - c.a = (int*) malloc(sizeof(int)*NUM); + c.a = (int *)malloc(sizeof(int) * NUM); for (int i = 0; i < NUM; i++) { c.a[i] = 1; } - #pragma omp target data map(mapper(id),tofrom: c) +#pragma omp target data map(mapper(id), tofrom : c) { - #pragma omp target teams distribute parallel for +#pragma omp target teams distribute parallel for for (int i = 0; i < NUM; i++) { ++c.a[i]; } @@ -33,4 +33,3 @@ int main() { printf("Sum = %d\n", sum); return 0; } - 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 index b714bca84dcb..f5fad8b8fe33 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_target_data_enter_exit.cpp @@ -10,20 +10,20 @@ public: int *a; }; -#pragma omp declare mapper(id: C s) map(s.a[0:NUM]) +#pragma omp declare mapper(id : C s) map(s.a[0 : NUM]) int main() { C c; - c.a = (int*) malloc(sizeof(int)*NUM); + c.a = (int *)malloc(sizeof(int) * NUM); for (int i = 0; i < NUM; i++) { c.a[i] = 1; } - #pragma omp target enter data map(mapper(id),to: c) - #pragma omp target teams distribute parallel for +#pragma omp target enter data map(mapper(id), to : c) +#pragma omp target teams distribute parallel for for (int i = 0; i < NUM; i++) { ++c.a[i]; } - #pragma omp target exit data map(mapper(id),from: c) +#pragma omp target exit data map(mapper(id), from : c) int sum = 0; for (int i = 0; i < NUM; i++) { sum += c.a[i]; @@ -32,4 +32,3 @@ int main() { printf("Sum = %d\n", sum); return 0; } - diff --git a/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp b/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp index 6aa23d20632f..fe4597b76908 100644 --- a/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp +++ b/openmp/libomptarget/test/mapping/declare_mapper_target_update.cpp @@ -10,21 +10,21 @@ public: int *a; }; -#pragma omp declare mapper(id: C s) map(s.a[0:NUM]) +#pragma omp declare mapper(id : C s) map(s.a[0 : NUM]) int main() { C c; int sum = 0; - c.a = (int*) malloc(sizeof(int)*NUM); + c.a = (int *)malloc(sizeof(int) * NUM); for (int i = 0; i < NUM; i++) { c.a[i] = 1; } - #pragma omp target enter data map(mapper(id),alloc: c) - #pragma omp target teams distribute parallel for +#pragma omp target enter data map(mapper(id), alloc : c) +#pragma omp target teams distribute parallel for for (int i = 0; i < NUM; i++) { c.a[i] = 0; } - #pragma omp target update from(mapper(id): c) +#pragma omp target update from(mapper(id) : c) for (int i = 0; i < NUM; i++) { sum += c.a[i]; } @@ -33,8 +33,8 @@ int main() { for (int i = 0; i < NUM; i++) { c.a[i] = 1; } - #pragma omp target update to(mapper(id): c) - #pragma omp target teams distribute parallel for +#pragma omp target update to(mapper(id) : c) +#pragma omp target teams distribute parallel for for (int i = 0; i < NUM; i++) { ++c.a[i]; } @@ -44,14 +44,13 @@ int main() { } // CHECK: Sum (after update to) = 1024 printf("Sum (after update to) = %d\n", sum); - #pragma omp target update from(mapper(id): c) +#pragma omp target update from(mapper(id) : c) sum = 0; for (int i = 0; i < NUM; i++) { sum += c.a[i]; } // CHECK: Sum (after second update from) = 2048 printf("Sum (after second update from) = %d\n", sum); - #pragma omp target exit data map(mapper(id),delete: c) +#pragma omp target exit data map(mapper(id), delete : c) return 0; } - diff --git a/openmp/libomptarget/test/mapping/delete_inf_refcount.c b/openmp/libomptarget/test/mapping/delete_inf_refcount.c index ca5980ff5170..f6016c19b0ff 100644 --- a/openmp/libomptarget/test/mapping/delete_inf_refcount.c +++ b/openmp/libomptarget/test/mapping/delete_inf_refcount.c @@ -1,7 +1,7 @@ // RUN: %libomptarget-compile-run-and-check-generic -#include <stdio.h> #include <omp.h> +#include <stdio.h> #pragma omp declare target int isHost; @@ -10,7 +10,7 @@ int isHost; int main(void) { isHost = -1; -#pragma omp target enter data map(to: isHost) +#pragma omp target enter data map(to : isHost) #pragma omp target { isHost = omp_is_initial_device(); } @@ -20,7 +20,7 @@ int main(void) { printf("Runtime error, isHost=%d\n", isHost); } -#pragma omp target exit data map(delete: isHost) +#pragma omp target exit data map(delete : isHost) // CHECK: Target region executed on the device printf("Target region executed on the %s\n", isHost ? "host" : "device"); diff --git a/openmp/libomptarget/test/mapping/device_ptr_update.c b/openmp/libomptarget/test/mapping/device_ptr_update.c index e3dd12f3d434..c6719d2285a2 100644 --- a/openmp/libomptarget/test/mapping/device_ptr_update.c +++ b/openmp/libomptarget/test/mapping/device_ptr_update.c @@ -15,17 +15,17 @@ int main(void) { s1.p = A; - // DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^ ]+]]{{\]}} - #pragma omp target enter data map(alloc : s1.p [0:10]) +// DEBUG: Update pointer ([[DEV_PTR:0x[^ ]+]]) -> {{\[}}[[DEV_OBJ_A:0x[^ ]+]]{{\]}} +#pragma omp target enter data map(alloc : s1.p[0 : 10]) - // DEBUG-NOT: Update pointer ([[DEV_PTR]]) -> {{\[}}[[DEV_OBJ_A]]{{\]}} - #pragma omp target map(alloc : s1.p [0:10]) +// DEBUG-NOT: Update pointer ([[DEV_PTR]]) -> {{\[}}[[DEV_OBJ_A]]{{\]}} +#pragma omp target map(alloc : s1.p[0 : 10]) { for (int i = 0; i < 10; ++i) s1.p[i] = i; } - #pragma omp target exit data map(from : s1.p [0:10]) +#pragma omp target exit data map(from : s1.p[0 : 10]) int fail_A = 0; for (int i = 0; i < 10; ++i) { diff --git a/openmp/libomptarget/test/mapping/lambda_by_value.cpp b/openmp/libomptarget/test/mapping/lambda_by_value.cpp index d922292e8595..22de281174a5 100644 --- a/openmp/libomptarget/test/mapping/lambda_by_value.cpp +++ b/openmp/libomptarget/test/mapping/lambda_by_value.cpp @@ -4,8 +4,8 @@ // XFAIL: amdgcn-amd-amdhsa // XFAIL: amdgcn-amd-amdhsa-LTO -#include <stdio.h> #include <stdint.h> +#include <stdio.h> // CHECK: before: [[V1:111]] [[V2:222]] [[PX:0x[^ ]+]] [[PY:0x[^ ]+]] // CHECK: lambda: [[V1]] [[V2]] [[PX_TGT:0x[^ ]+]] 0x{{.*}} @@ -25,14 +25,13 @@ int main() { printf("before: %d %ld %p %p\n", x[1], y[1], &x[0], &y[0]); - intptr_t xp = (intptr_t) &x[0]; + intptr_t xp = (intptr_t)&x[0]; #pragma omp target firstprivate(xp) { lambda(); - printf("tgt : %d %p %d\n", x[1], &x[0], (&x[0] != (int*) xp)); + printf("tgt : %d %p %d\n", x[1], &x[0], (&x[0] != (int *)xp)); } printf("out : %d %ld %p %p\n", x[1], y[1], &x[0], &y[0]); return 0; } - diff --git a/openmp/libomptarget/test/mapping/lambda_mapping.cpp b/openmp/libomptarget/test/mapping/lambda_mapping.cpp index c085d9fc110a..7881880c3761 100644 --- a/openmp/libomptarget/test/mapping/lambda_mapping.cpp +++ b/openmp/libomptarget/test/mapping/lambda_mapping.cpp @@ -29,7 +29,7 @@ int main() { C[I] = -9; } -#pragma omp target data map(tofrom : C [0:N]) map(to : A [0:N], B [0:N]) +#pragma omp target data map(tofrom : C[0 : N]) map(to : A[0 : N], B[0 : N]) { forall(0, N, [&](int I) { C[I] += A[I] + B[I]; }); } diff --git a/openmp/libomptarget/test/mapping/ompx_hold/omp_target_disassociate_ptr.c b/openmp/libomptarget/test/mapping/ompx_hold/omp_target_disassociate_ptr.c index 6bd0e9199177..ea4a1ffc979a 100644 --- a/openmp/libomptarget/test/mapping/ompx_hold/omp_target_disassociate_ptr.c +++ b/openmp/libomptarget/test/mapping/ompx_hold/omp_target_disassociate_ptr.c @@ -12,9 +12,9 @@ // RUN: %not %libomptarget-run-generic 1 2>&1 | %fcheck-generic // RUN: %not %libomptarget-run-generic inf 2>&1 | %fcheck-generic +#include <limits.h> #include <omp.h> #include <stdio.h> -#include <limits.h> #include <string.h> int main(int argc, char *argv[]) { @@ -44,16 +44,16 @@ int main(int argc, char *argv[]) { } } else { for (int I = 0; I < DynRef; ++I) { - #pragma omp target enter data map(alloc: X) +#pragma omp target enter data map(alloc : X) } } // Disassociate while hold reference count > 0. int Status = 0; - #pragma omp target data map(ompx_hold,alloc: X) +#pragma omp target data map(ompx_hold, alloc : X) #if HOLD_MORE - #pragma omp target data map(ompx_hold,alloc: X) - #pragma omp target data map(ompx_hold,alloc: X) +#pragma omp target data map(ompx_hold, alloc : X) +#pragma omp target data map(ompx_hold, alloc : X) #endif { // CHECK: Libomptarget error: Trying to disassociate a pointer with a diff --git a/openmp/libomptarget/test/mapping/ompx_hold/struct.c b/openmp/libomptarget/test/mapping/ompx_hold/struct.c index cc2910b62f64..63ba44a7bc34 100644 --- a/openmp/libomptarget/test/mapping/ompx_hold/struct.c +++ b/openmp/libomptarget/test/mapping/ompx_hold/struct.c @@ -9,18 +9,19 @@ #include <stdio.h> #define CHECK_PRESENCE(Var1, Var2, Var3) \ - printf(" presence of %s, %s, %s: %d, %d, %d\n", \ - #Var1, #Var2, #Var3, \ + printf(" presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3, \ omp_target_is_present(&(Var1), omp_get_default_device()), \ omp_target_is_present(&(Var2), omp_get_default_device()), \ omp_target_is_present(&(Var3), omp_get_default_device())) #define CHECK_VALUES(Var1, Var2) \ - printf(" values of %s, %s: %d, %d\n", \ - #Var1, #Var2, (Var1), (Var2)) + printf(" values of %s, %s: %d, %d\n", #Var1, #Var2, (Var1), (Var2)) int main() { - struct S { int i; int j; } s; + struct S { + int i; + int j; + } s; // CHECK: presence of s, s.i, s.j: 0, 0, 0 CHECK_PRESENCE(s, s.i, s.j); @@ -32,17 +33,17 @@ int main() { printf("check: ompx_hold only on first member\n"); s.i = 20; s.j = 30; - #pragma omp target data map(tofrom: s) map(ompx_hold,tofrom: s.i) \ - map(tofrom: s.j) +#pragma omp target data map(tofrom : s) map(ompx_hold, tofrom : s.i) \ + map(tofrom : s.j) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(tofrom: s) +#pragma omp target map(tofrom : s) { s.i = 21; s.j = 31; } - #pragma omp target exit data map(delete: s, s.i) +#pragma omp target exit data map(delete : s, s.i) // ompx_hold on s.i applies to all of s. // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 // CHECK-NEXT: values of s.i, s.j: 20, 30 @@ -59,17 +60,17 @@ int main() { printf("check: ompx_hold only on last member\n"); s.i = 20; s.j = 30; - #pragma omp target data map(tofrom: s) map(tofrom: s.i) \ - map(ompx_hold,tofrom: s.j) +#pragma omp target data map(tofrom : s) map(tofrom : s.i) \ + map(ompx_hold, tofrom : s.j) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(tofrom: s) +#pragma omp target map(tofrom : s) { s.i = 21; s.j = 31; } - #pragma omp target exit data map(delete: s, s.i) +#pragma omp target exit data map(delete : s, s.i) // ompx_hold on s.j applies to all of s. // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 // CHECK-NEXT: values of s.i, s.j: 20, 30 @@ -86,17 +87,17 @@ int main() { printf("check: ompx_hold only on struct\n"); s.i = 20; s.j = 30; - #pragma omp target data map(ompx_hold,tofrom: s) map(tofrom: s.i) \ - map(tofrom: s.j) +#pragma omp target data map(ompx_hold, tofrom : s) map(tofrom : s.i) \ + map(tofrom : s.j) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(tofrom: s) +#pragma omp target map(tofrom : s) { s.i = 21; s.j = 31; } - #pragma omp target exit data map(delete: s, s.i) +#pragma omp target exit data map(delete : s, s.i) // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 // CHECK-NEXT: values of s.i, s.j: 20, 30 CHECK_PRESENCE(s, s.i, s.j); @@ -115,12 +116,12 @@ int main() { printf("check: parent DynRefCount=1 is not sufficient for transfer\n"); s.i = 20; s.j = 30; - #pragma omp target data map(ompx_hold, tofrom: s) - #pragma omp target data map(ompx_hold, tofrom: s) +#pragma omp target data map(ompx_hold, tofrom : s) +#pragma omp target data map(ompx_hold, tofrom : s) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(from: s.i, s.j) +#pragma omp target map(from : s.i, s.j) { s.i = 21; s.j = 31; @@ -129,7 +130,7 @@ int main() { // CHECK-NEXT: values of s.i, s.j: 20, 30 CHECK_PRESENCE(s, s.i, s.j); CHECK_VALUES(s.i, s.j); - #pragma omp target map(to: s.i, s.j) +#pragma omp target map(to : s.i, s.j) { // No transfer here even though parent's DynRefCount=1. // CHECK-NEXT: values of s.i, s.j: 21, 31 CHECK_VALUES(s.i, s.j); @@ -145,12 +146,12 @@ int main() { printf("check: parent HoldRefCount=1 is not sufficient for transfer\n"); s.i = 20; s.j = 30; - #pragma omp target data map(tofrom: s) - #pragma omp target data map(tofrom: s) +#pragma omp target data map(tofrom : s) +#pragma omp target data map(tofrom : s) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(ompx_hold, from: s.i, s.j) +#pragma omp target map(ompx_hold, from : s.i, s.j) { s.i = 21; s.j = 31; @@ -159,7 +160,7 @@ int main() { // CHECK-NEXT: values of s.i, s.j: 20, 30 CHECK_PRESENCE(s, s.i, s.j); CHECK_VALUES(s.i, s.j); - #pragma omp target map(ompx_hold, to: s.i, s.j) +#pragma omp target map(ompx_hold, to : s.i, s.j) { // No transfer here even though parent's HoldRefCount=1. // CHECK-NEXT: values of s.i, s.j: 21, 31 CHECK_VALUES(s.i, s.j); @@ -181,16 +182,16 @@ int main() { printf("check: parent TotalRefCount=1 is not sufficient for transfer\n"); s.i = 20; s.j = 30; - #pragma omp target data map(ompx_hold, tofrom: s) +#pragma omp target data map(ompx_hold, tofrom : s) { // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 CHECK_PRESENCE(s, s.i, s.j); - #pragma omp target map(ompx_hold, tofrom: s.i, s.j) +#pragma omp target map(ompx_hold, tofrom : s.i, s.j) { s.i = 21; s.j = 31; } - #pragma omp target exit data map(from: s.i, s.j) +#pragma omp target exit data map(from : s.i, s.j) // No transfer here even though parent's TotalRefCount=1. // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1 // CHECK-NEXT: values of s.i, s.j: 20, 30 diff --git a/openmp/libomptarget/test/mapping/ompx_hold/target-data.c b/openmp/libomptarget/test/mapping/ompx_hold/target-data.c index 154eb0358391..dd784591344b 100644 --- a/openmp/libomptarget/test/mapping/ompx_hold/target-data.c +++ b/openmp/libomptarget/test/mapping/ompx_hold/target-data.c @@ -5,8 +5,7 @@ #include <stdio.h> #define CHECK_PRESENCE(Var1, Var2, Var3) \ - printf(" presence of %s, %s, %s: %d, %d, %d\n", \ - #Var1, #Var2, #Var3, \ + printf(" presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3, \ omp_target_is_present(&Var1, omp_get_default_device()), \ omp_target_is_present(&Var2, omp_get_default_device()), \ omp_target_is_present(&Var3, omp_get_default_device())) @@ -22,11 +21,11 @@ int main() { // CHECK-NEXT: structured{{.*}} printf(" structured dec of dyn\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -39,24 +38,24 @@ int main() { // CHECK-NEXT: dynamic{{.*}} printf(" dynamic dec/reset of dyn\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) +#pragma omp target exit data map(from : m) map(release : r) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); } // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); } @@ -71,25 +70,24 @@ int main() { // CHECK-NEXT: dynamic{{.*}} printf(" dynamic dec/reset of dyn\n"); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) +#pragma omp target exit data map(from : m) map(release : r) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); } // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); } @@ -102,20 +100,19 @@ int main() { // CHECK-NEXT: structured{{.*}} printf(" structured dec of dyn\n"); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -134,31 +131,30 @@ int main() { // CHECK-NEXT: dynamic{{.*}} printf(" dynamic dec/reset of dyn\n"); - #pragma omp target enter data map(to: m) map(alloc: r, d) +#pragma omp target enter data map(to : m) map(alloc : r, d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target enter data map(to: m) map(alloc: r, d) +#pragma omp target enter data map(to : m) map(alloc : r, d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) +#pragma omp target exit data map(from : m) map(release : r) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); } // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); } @@ -171,21 +167,19 @@ int main() { // CHECK-NEXT: structured{{.*}} printf(" structured dec of dyn\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -204,18 +198,17 @@ int main() { // CHECK-NEXT: dynamic{{.*}} printf(" dynamic dec/reset of dyn\n"); - #pragma omp target enter data map(to: m) map(alloc: r, d) +#pragma omp target enter data map(to : m) map(alloc : r, d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target enter data map(to: m) map(alloc: r, d) +#pragma omp target enter data map(to : m) map(alloc : r, d) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -225,10 +218,10 @@ int main() { } // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) +#pragma omp target exit data map(from : m) map(release : r) // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target exit data map(from: m) map(release: r) map(delete: d) +#pragma omp target exit data map(from : m) map(release : r) map(delete : d) // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); diff --git a/openmp/libomptarget/test/mapping/ompx_hold/target.c b/openmp/libomptarget/test/mapping/ompx_hold/target.c index 614970684aa3..40fd14c6bf00 100644 --- a/openmp/libomptarget/test/mapping/ompx_hold/target.c +++ b/openmp/libomptarget/test/mapping/ompx_hold/target.c @@ -5,8 +5,7 @@ #include <stdio.h> #define CHECK_PRESENCE(Var1, Var2, Var3) \ - printf(" presence of %s, %s, %s: %d, %d, %d\n", \ - #Var1, #Var2, #Var3, \ + printf(" presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3, \ omp_target_is_present(&Var1, omp_get_default_device()), \ omp_target_is_present(&Var2, omp_get_default_device()), \ omp_target_is_present(&Var3, omp_get_default_device())) @@ -22,18 +21,18 @@ int main() { // CHECK-NEXT: once printf(" once\n"); - #pragma omp target map(tofrom: m) map(alloc: r, d) +#pragma omp target map(tofrom : m) map(alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); // CHECK-NEXT: twice printf(" twice\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(tofrom: m) map(alloc: r, d) +#pragma omp target map(tofrom : m) map(alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -47,18 +46,18 @@ int main() { // CHECK-NEXT: once printf(" once\n"); - #pragma omp target map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 0, 0, 0 CHECK_PRESENCE(m, r, d); // CHECK-NEXT: twice printf(" twice\n"); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -72,11 +71,11 @@ int main() { // CHECK-NEXT: once each printf(" once each\n"); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(tofrom: m) map(alloc: r, d) +#pragma omp target map(tofrom : m) map(alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -86,20 +85,19 @@ int main() { // CHECK-NEXT: twice each printf(" twice each\n"); - #pragma omp target data map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(tofrom: m) map(alloc: r, d) +#pragma omp target map(tofrom : m) map(alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -119,11 +117,11 @@ int main() { // CHECK-NEXT: once each printf(" once each\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); @@ -133,20 +131,19 @@ int main() { // CHECK-NEXT: twice each printf(" twice each\n"); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(tofrom: m) map(alloc: r, d) +#pragma omp target data map(tofrom : m) map(alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target data map(ompx_hold, tofrom: m) \ - map(ompx_hold, alloc: r, d) +#pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) { // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); - #pragma omp target map(ompx_hold, tofrom: m) map(ompx_hold, alloc: r, d) +#pragma omp target map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d) ; // CHECK-NEXT: presence of m, r, d: 1, 1, 1 CHECK_PRESENCE(m, r, d); diff --git a/openmp/libomptarget/test/mapping/pr38704.c b/openmp/libomptarget/test/mapping/pr38704.c index 28cab8a65515..72b4a171ccea 100644 --- a/openmp/libomptarget/test/mapping/pr38704.c +++ b/openmp/libomptarget/test/mapping/pr38704.c @@ -19,14 +19,14 @@ int main(int argc, char *argv[]) { s2.ptr1 = malloc(sizeof(int)); s2.ptr2 = malloc(2 * sizeof(int)); -#pragma omp target enter data map(to: s2.ptr2[0:1]) -#pragma omp target map(s.ptr1[0:1], s.ptr2[0:2]) +#pragma omp target enter data map(to : s2.ptr2[0 : 1]) +#pragma omp target map(s.ptr1[0 : 1], s.ptr2[0 : 2]) { s.ptr1[0] = 1; s.ptr2[0] = 2; s.ptr2[1] = 3; } -#pragma omp target exit data map(from: s2.ptr1[0:1], s2.ptr2[0:1]) +#pragma omp target exit data map(from : s2.ptr1[0 : 1], s2.ptr2[0 : 1]) // CHECK: s.ptr1[0] = 1 // CHECK: s.ptr2[0] = 2 diff --git a/openmp/libomptarget/test/mapping/present/target.c b/openmp/libomptarget/test/mapping/present/target.c index 67dc6d3e9998..cf597a00da32 100644 --- a/openmp/libomptarget/test/mapping/present/target.c +++ b/openmp/libomptarget/test/mapping/present/target.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -12,8 +11,8 @@ int main() { fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: i) -#pragma omp target map(present, alloc: i) +#pragma omp target data map(alloc : i) +#pragma omp target map(present, alloc : i) ; // CHECK: i is present @@ -24,7 +23,7 @@ int main() { // CHECK: Libomptarget error: Call to targetDataBegin failed, abort target. // CHECK: Libomptarget error: Failed to process data before launching the kernel. // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target map(present, alloc: i) +#pragma omp target map(present, alloc : i) ; // CHECK-NOT: i is present diff --git a/openmp/libomptarget/test/mapping/present/target_array_extension.c b/openmp/libomptarget/test/mapping/present/target_array_extension.c index 5545f9052b9a..aa50e09748db 100644 --- a/openmp/libomptarget/test/mapping/present/target_array_extension.c +++ b/openmp/libomptarget/test/mapping/present/target_array_extension.c @@ -16,35 +16,38 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <stdio.h> #define BEFORE 0 -#define AFTER 1 +#define AFTER 1 #define SIZE 100 #if EXTENDS == BEFORE -# define SMALL_BEG (SIZE-2) -# define SMALL_END SIZE -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG (SIZE - 2) +#define SMALL_END SIZE +#define LARGE_BEG 0 +#define LARGE_END SIZE #elif EXTENDS == AFTER -# define SMALL_BEG 0 -# define SMALL_END 2 -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG 0 +#define SMALL_END 2 +#define LARGE_BEG 0 +#define LARGE_END SIZE #else -# error EXTENDS undefined +#error EXTENDS undefined #endif -#define SMALL_SIZE (SMALL_END-SMALL_BEG) -#define LARGE_SIZE (LARGE_END-LARGE_BEG) +#define SMALL_SIZE (SMALL_END - SMALL_BEG) +#define LARGE_SIZE (LARGE_END - LARGE_BEG) -#define SMALL SMALL_BEG:SMALL_SIZE -#define LARGE LARGE_BEG:LARGE_SIZE +#define SMALL \ + SMALL_BEG: \ + SMALL_SIZE +#define LARGE \ + LARGE_BEG: \ + LARGE_SIZE int main() { int arr[SIZE]; @@ -58,9 +61,9 @@ int main() { LARGE_SIZE * sizeof arr[0]); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[LARGE]) +#pragma omp target data map(alloc : arr[LARGE]) { -#pragma omp target map(present, tofrom: arr[SMALL]) +#pragma omp target map(present, tofrom : arr[SMALL]) ; } @@ -73,9 +76,9 @@ int main() { // CHECK: Libomptarget error: Call to targetDataBegin failed, abort target. // CHECK: Libomptarget error: Failed to process data before launching the kernel. // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target data map(alloc: arr[SMALL]) +#pragma omp target data map(alloc : arr[SMALL]) { -#pragma omp target map(present, tofrom: arr[LARGE]) +#pragma omp target map(present, tofrom : arr[LARGE]) ; } diff --git a/openmp/libomptarget/test/mapping/present/target_data.c b/openmp/libomptarget/test/mapping/present/target_data.c index d58f7c8dc04a..1d68292059b4 100644 --- a/openmp/libomptarget/test/mapping/present/target_data.c +++ b/openmp/libomptarget/test/mapping/present/target_data.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -12,8 +11,8 @@ int main() { fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: i) -#pragma omp target data map(present, alloc: i) +#pragma omp target data map(alloc : i) +#pragma omp target data map(present, alloc : i) ; // CHECK: i is present @@ -21,7 +20,7 @@ int main() { // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target data map(present, alloc: i) +#pragma omp target data map(present, alloc : i) ; // CHECK-NOT: i is present diff --git a/openmp/libomptarget/test/mapping/present/target_data_array_extension.c b/openmp/libomptarget/test/mapping/present/target_data_array_extension.c index 9964e006383d..9043596006a3 100644 --- a/openmp/libomptarget/test/mapping/present/target_data_array_extension.c +++ b/openmp/libomptarget/test/mapping/present/target_data_array_extension.c @@ -16,35 +16,38 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <stdio.h> #define BEFORE 0 -#define AFTER 1 +#define AFTER 1 #define SIZE 100 #if EXTENDS == BEFORE -# define SMALL_BEG (SIZE-2) -# define SMALL_END SIZE -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG (SIZE - 2) +#define SMALL_END SIZE +#define LARGE_BEG 0 +#define LARGE_END SIZE #elif EXTENDS == AFTER -# define SMALL_BEG 0 -# define SMALL_END 2 -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG 0 +#define SMALL_END 2 +#define LARGE_BEG 0 +#define LARGE_END SIZE #else -# error EXTENDS undefined +#error EXTENDS undefined #endif -#define SMALL_SIZE (SMALL_END-SMALL_BEG) -#define LARGE_SIZE (LARGE_END-LARGE_BEG) +#define SMALL_SIZE (SMALL_END - SMALL_BEG) +#define LARGE_SIZE (LARGE_END - LARGE_BEG) -#define SMALL SMALL_BEG:SMALL_SIZE -#define LARGE LARGE_BEG:LARGE_SIZE +#define SMALL \ + SMALL_BEG: \ + SMALL_SIZE +#define LARGE \ + LARGE_BEG: \ + LARGE_SIZE int main() { int arr[SIZE]; @@ -58,9 +61,9 @@ int main() { LARGE_SIZE * sizeof arr[0]); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[LARGE]) +#pragma omp target data map(alloc : arr[LARGE]) { -#pragma omp target data map(present, tofrom: arr[SMALL]) +#pragma omp target data map(present, tofrom : arr[SMALL]) ; } @@ -71,9 +74,9 @@ int main() { // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes) // CHECK: Libomptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier). // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target data map(alloc: arr[SMALL]) +#pragma omp target data map(alloc : arr[SMALL]) { -#pragma omp target data map(present, tofrom: arr[LARGE]) +#pragma omp target data map(present, tofrom : arr[LARGE]) ; } diff --git a/openmp/libomptarget/test/mapping/present/target_data_at_exit.c b/openmp/libomptarget/test/mapping/present/target_data_at_exit.c index 36ce90871454..fd48cf472505 100644 --- a/openmp/libomptarget/test/mapping/present/target_data_at_exit.c +++ b/openmp/libomptarget/test/mapping/present/target_data_at_exit.c @@ -2,19 +2,18 @@ // RUN: %libomptarget-run-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { int i; -#pragma omp target enter data map(alloc:i) +#pragma omp target enter data map(alloc : i) // i isn't present at the end of the target data region, but the "present" // modifier is only checked at the beginning of a region. -#pragma omp target data map(present, alloc: i) +#pragma omp target data map(present, alloc : i) { -#pragma omp target exit data map(delete:i) +#pragma omp target exit data map(delete : i) } // CHECK-NOT: Libomptarget diff --git a/openmp/libomptarget/test/mapping/present/target_enter_data.c b/openmp/libomptarget/test/mapping/present/target_enter_data.c index 2bdbd308ca79..0bb9ac9f02a4 100644 --- a/openmp/libomptarget/test/mapping/present/target_enter_data.c +++ b/openmp/libomptarget/test/mapping/present/target_enter_data.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -12,9 +11,9 @@ int main() { fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); // CHECK-NOT: Libomptarget -#pragma omp target enter data map(alloc: i) -#pragma omp target enter data map(present, alloc: i) -#pragma omp target exit data map(delete: i) +#pragma omp target enter data map(alloc : i) +#pragma omp target enter data map(present, alloc : i) +#pragma omp target exit data map(delete : i) // CHECK: i is present fprintf(stderr, "i is present\n"); @@ -22,7 +21,7 @@ int main() { // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) // CHECK: Libomptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier). // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target enter data map(present, alloc: i) +#pragma omp target enter data map(present, alloc : i) // CHECK-NOT: i is present fprintf(stderr, "i is present\n"); diff --git a/openmp/libomptarget/test/mapping/present/target_exit_data_delete.c b/openmp/libomptarget/test/mapping/present/target_exit_data_delete.c index 7585ce8d7978..e73b8e6086c6 100644 --- a/openmp/libomptarget/test/mapping/present/target_exit_data_delete.c +++ b/openmp/libomptarget/test/mapping/present/target_exit_data_delete.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -11,16 +10,16 @@ int main() { // CHECK: addr=0x[[#%x,HOST_ADDR:]], size=[[#%u,SIZE:]] fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); - // CHECK-NOT: Libomptarget - #pragma omp target enter data map(alloc: i) - #pragma omp target exit data map(present, delete: i) +// CHECK-NOT: Libomptarget +#pragma omp target enter data map(alloc : i) +#pragma omp target exit data map(present, delete : i) // CHECK: i was present fprintf(stderr, "i was present\n"); - // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) - // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory - #pragma omp target exit data map(present, delete: i) +// CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) +// CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory +#pragma omp target exit data map(present, delete : i) // CHECK-NOT: i was present fprintf(stderr, "i was present\n"); diff --git a/openmp/libomptarget/test/mapping/present/target_exit_data_release.c b/openmp/libomptarget/test/mapping/present/target_exit_data_release.c index 6b32fbe74418..4110f0f3f83b 100644 --- a/openmp/libomptarget/test/mapping/present/target_exit_data_release.c +++ b/openmp/libomptarget/test/mapping/present/target_exit_data_release.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -11,16 +10,16 @@ int main() { // CHECK: addr=0x[[#%x,HOST_ADDR:]], size=[[#%u,SIZE:]] fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); - // CHECK-NOT: Libomptarget - #pragma omp target enter data map(alloc: i) - #pragma omp target exit data map(present, release: i) +// CHECK-NOT: Libomptarget +#pragma omp target enter data map(alloc : i) +#pragma omp target exit data map(present, release : i) // CHECK: i was present fprintf(stderr, "i was present\n"); - // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) - // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory - #pragma omp target exit data map(present, release: i) +// CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) +// CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory +#pragma omp target exit data map(present, release : i) // CHECK-NOT: i was present fprintf(stderr, "i was present\n"); diff --git a/openmp/libomptarget/test/mapping/present/target_update.c b/openmp/libomptarget/test/mapping/present/target_update.c index 3ec126f02050..7d7d5f85e1f3 100644 --- a/openmp/libomptarget/test/mapping/present/target_update.c +++ b/openmp/libomptarget/test/mapping/present/target_update.c @@ -16,7 +16,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -26,16 +25,16 @@ int main() { fprintf(stderr, "addr=%p, size=%ld\n", &i, sizeof i); // CHECK-NOT: Libomptarget -#pragma omp target enter data map(alloc: i) -#pragma omp target update CLAUSE(present: i) -#pragma omp target exit data map(delete: i) +#pragma omp target enter data map(alloc : i) +#pragma omp target update CLAUSE(present : i) +#pragma omp target exit data map(delete : i) // CHECK: i is present fprintf(stderr, "i is present\n"); // CHECK: Libomptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target update CLAUSE(present: i) +#pragma omp target update CLAUSE(present : i) // CHECK-NOT: i is present fprintf(stderr, "i is present\n"); diff --git a/openmp/libomptarget/test/mapping/present/target_update_array_extension.c b/openmp/libomptarget/test/mapping/present/target_update_array_extension.c index 9ab1c78d49bf..e1a60339cc7e 100644 --- a/openmp/libomptarget/test/mapping/present/target_update_array_extension.c +++ b/openmp/libomptarget/test/mapping/present/target_update_array_extension.c @@ -34,22 +34,21 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <stdio.h> #define BEFORE 0 -#define AFTER 1 +#define AFTER 1 #if EXTENDS == BEFORE -# define SMALL 2:3 -# define LARGE 0:5 +#define SMALL 2 : 3 +#define LARGE 0 : 5 #elif EXTENDS == AFTER -# define SMALL 0:3 -# define LARGE 0:5 +#define SMALL 0 : 3 +#define LARGE 0 : 5 #else -# error EXTENDS undefined +#error EXTENDS undefined #endif int main() { @@ -59,9 +58,9 @@ int main() { fprintf(stderr, "addr=%p, size=%ld\n", arr, sizeof arr); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[LARGE]) +#pragma omp target data map(alloc : arr[LARGE]) { -#pragma omp target update CLAUSE(present: arr[SMALL]) +#pragma omp target update CLAUSE(present : arr[SMALL]) } // CHECK: arr is present @@ -69,9 +68,9 @@ int main() { // CHECK: Libomptarget message: device mapping required by 'present' motion modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target data map(alloc: arr[SMALL]) +#pragma omp target data map(alloc : arr[SMALL]) { -#pragma omp target update CLAUSE(present: arr[LARGE]) +#pragma omp target update CLAUSE(present : arr[LARGE]) } // CHECK-NOT: arr is present diff --git a/openmp/libomptarget/test/mapping/present/unified_shared_memory.c b/openmp/libomptarget/test/mapping/present/unified_shared_memory.c index 2493685607cd..c5a2fe42928f 100644 --- a/openmp/libomptarget/test/mapping/present/unified_shared_memory.c +++ b/openmp/libomptarget/test/mapping/present/unified_shared_memory.c @@ -4,7 +4,6 @@ // REQUIRES: unified_shared_memory - #include <stdio.h> // The runtime considers unified shared memory to be always present. @@ -14,15 +13,15 @@ int main() { int i; // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: i) -#pragma omp target map(present, alloc: i) +#pragma omp target data map(alloc : i) +#pragma omp target map(present, alloc : i) ; // CHECK: i is present fprintf(stderr, "i is present\n"); // CHECK-NOT: Libomptarget -#pragma omp target map(present, alloc: i) +#pragma omp target map(present, alloc : i) ; // CHECK: is present diff --git a/openmp/libomptarget/test/mapping/present/zero_length_array_section.c b/openmp/libomptarget/test/mapping/present/zero_length_array_section.c index f43053e13716..727ee4bb7cdb 100644 --- a/openmp/libomptarget/test/mapping/present/zero_length_array_section.c +++ b/openmp/libomptarget/test/mapping/present/zero_length_array_section.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -12,8 +11,8 @@ int main() { fprintf(stderr, "addr=%p\n", arr); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[0:5]) -#pragma omp target map(present, alloc: arr[0:0]) +#pragma omp target data map(alloc : arr[0 : 5]) +#pragma omp target map(present, alloc : arr[0 : 0]) ; // CHECK: arr is present @@ -26,8 +25,8 @@ int main() { // CHECK: Libomptarget error: Call to targetDataBegin failed, abort target. // CHECK: Libomptarget error: Failed to process data before launching the kernel. // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target data map(alloc: arr[0:0]) -#pragma omp target map(present, alloc: arr[0:0]) +#pragma omp target data map(alloc : arr[0 : 0]) +#pragma omp target map(present, alloc : arr[0 : 0]) ; // CHECK-NOT: arr is present diff --git a/openmp/libomptarget/test/mapping/present/zero_length_array_section_exit.c b/openmp/libomptarget/test/mapping/present/zero_length_array_section_exit.c index bc441bf98f34..86beb03c2bf8 100644 --- a/openmp/libomptarget/test/mapping/present/zero_length_array_section_exit.c +++ b/openmp/libomptarget/test/mapping/present/zero_length_array_section_exit.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-fail-generic 2>&1 \ // RUN: | %fcheck-generic - #include <stdio.h> int main() { @@ -12,8 +11,8 @@ int main() { fprintf(stderr, "addr=%p\n", arr); // CHECK-NOT: Libomptarget -#pragma omp target enter data map(alloc: arr[0:5]) -#pragma omp target exit data map(present, release: arr[0:0]) +#pragma omp target enter data map(alloc : arr[0 : 5]) +#pragma omp target exit data map(present, release : arr[0 : 0]) // CHECK: arr is present fprintf(stderr, "arr is present\n"); @@ -22,8 +21,8 @@ int main() { // // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] (0 bytes) // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory -#pragma omp target enter data map(alloc: arr[0:0]) -#pragma omp target exit data map(present, release: arr[0:0]) +#pragma omp target enter data map(alloc : arr[0 : 0]) +#pragma omp target exit data map(present, release : arr[0 : 0]) // CHECK-NOT: arr is present fprintf(stderr, "arr is present\n"); diff --git a/openmp/libomptarget/test/mapping/private_mapping.c b/openmp/libomptarget/test/mapping/private_mapping.c index 4bd1c37ae36a..6ed8cbf9e488 100644 --- a/openmp/libomptarget/test/mapping/private_mapping.c +++ b/openmp/libomptarget/test/mapping/private_mapping.c @@ -6,8 +6,7 @@ int main() { int data1[3] = {1}, data2[3] = {2}, data3[3] = {3}; int sum[16] = {0}; -#pragma omp target teams distribute parallel for map(tofrom \ - : sum) \ +#pragma omp target teams distribute parallel for map(tofrom : sum) \ firstprivate(data1, data2, data3) for (int i = 0; i < 16; ++i) { for (int j = 0; j < 3; ++j) { diff --git a/openmp/libomptarget/test/mapping/ptr_and_obj_motion.c b/openmp/libomptarget/test/mapping/ptr_and_obj_motion.c index a975ed63bc20..8fa2c9865b4a 100644 --- a/openmp/libomptarget/test/mapping/ptr_and_obj_motion.c +++ b/openmp/libomptarget/test/mapping/ptr_and_obj_motion.c @@ -9,7 +9,7 @@ typedef struct { } DV; void init(double vertexx[]) { - #pragma omp target map(vertexx[0:100]) +#pragma omp target map(vertexx[0 : 100]) { printf("In init: %lf, expected 100.0\n", vertexx[77]); vertexx[77] = 77.0; @@ -17,7 +17,7 @@ void init(double vertexx[]) { } void change(DV *dvptr) { - #pragma omp target map(dvptr->dataptr[0:100]) +#pragma omp target map(dvptr->dataptr[0 : 100]) { printf("In change: %lf, expected 77.0\n", dvptr->dataptr[77]); dvptr->dataptr[77] += 1.0; @@ -31,14 +31,13 @@ int main() { DV dv; dv.dataptr = &vertexx[0]; - #pragma omp target enter data map(to:vertexx[0:100]) +#pragma omp target enter data map(to : vertexx[0 : 100]) init(vertexx); change(&dv); - #pragma omp target exit data map(from:vertexx[0:100]) +#pragma omp target exit data map(from : vertexx[0 : 100]) // CHECK: Final: 78.0 printf("Final: %lf\n", vertexx[77]); } - diff --git a/openmp/libomptarget/test/mapping/reduction_implicit_map.cpp b/openmp/libomptarget/test/mapping/reduction_implicit_map.cpp index 429bb68c43e9..329aa43ddf3c 100644 --- a/openmp/libomptarget/test/mapping/reduction_implicit_map.cpp +++ b/openmp/libomptarget/test/mapping/reduction_implicit_map.cpp @@ -2,15 +2,13 @@ #include <stdio.h> -void sum(int* input, int size, int* output) -{ -#pragma omp target teams distribute parallel for reduction(+:output[0]) \ - map(to:input[0:size]) +void sum(int *input, int size, int *output) { +#pragma omp target teams distribute parallel for reduction(+ : output[0]) \ + map(to : input[0 : size]) for (int i = 0; i < size; i++) output[0] += input[i]; } -int main() -{ +int main() { const int size = 100; int *array = new int[size]; int result = 0; @@ -22,4 +20,3 @@ int main() delete[] array; return 0; } - diff --git a/openmp/libomptarget/test/mapping/target_data_array_extension_at_exit.c b/openmp/libomptarget/test/mapping/target_data_array_extension_at_exit.c index 11d53ba2a5e7..58020620e45c 100644 --- a/openmp/libomptarget/test/mapping/target_data_array_extension_at_exit.c +++ b/openmp/libomptarget/test/mapping/target_data_array_extension_at_exit.c @@ -16,32 +16,35 @@ // RUN: %libomptarget-run-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <stdio.h> #define BEFORE 0 -#define AFTER 1 +#define AFTER 1 #define SIZE 100 #if EXTENDS == BEFORE -# define SMALL_BEG (SIZE-2) -# define SMALL_END SIZE -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG (SIZE - 2) +#define SMALL_END SIZE +#define LARGE_BEG 0 +#define LARGE_END SIZE #elif EXTENDS == AFTER -# define SMALL_BEG 0 -# define SMALL_END 2 -# define LARGE_BEG 0 -# define LARGE_END SIZE +#define SMALL_BEG 0 +#define SMALL_END 2 +#define LARGE_BEG 0 +#define LARGE_END SIZE #else -# error EXTENDS undefined +#error EXTENDS undefined #endif -#define SMALL SMALL_BEG:(SMALL_END-SMALL_BEG) -#define LARGE LARGE_BEG:(LARGE_END-LARGE_BEG) +#define SMALL \ + SMALL_BEG: \ + (SMALL_END - SMALL_BEG) +#define LARGE \ + LARGE_BEG: \ + (LARGE_END - LARGE_BEG) void check_not_present() { int arr[SIZE]; @@ -54,11 +57,11 @@ void check_not_present() { // arr[LARGE] isn't (fully) present at the end of the target data region, so // the device-to-host transfer should not be performed, or it might fail. -#pragma omp target data map(tofrom: arr[LARGE]) +#pragma omp target data map(tofrom : arr[LARGE]) { -#pragma omp target exit data map(delete: arr[LARGE]) -#pragma omp target enter data map(alloc: arr[SMALL]) -#pragma omp target map(alloc: arr[SMALL]) +#pragma omp target exit data map(delete : arr[LARGE]) +#pragma omp target enter data map(alloc : arr[SMALL]) +#pragma omp target map(alloc : arr[SMALL]) for (int i = SMALL_BEG; i < SMALL_END; ++i) arr[i] = 88; } @@ -83,11 +86,11 @@ void check_is_present() { // arr[SMALL] is (fully) present at the end of the target data region, and the // device-to-host transfer should be performed only for it even though more // of the array is then present. -#pragma omp target data map(tofrom: arr[SMALL]) +#pragma omp target data map(tofrom : arr[SMALL]) { -#pragma omp target exit data map(delete: arr[SMALL]) -#pragma omp target enter data map(alloc: arr[LARGE]) -#pragma omp target map(alloc: arr[LARGE]) +#pragma omp target exit data map(delete : arr[SMALL]) +#pragma omp target enter data map(alloc : arr[LARGE]) +#pragma omp target map(alloc : arr[LARGE]) for (int i = LARGE_BEG; i < LARGE_END; ++i) arr[i] = 88; } diff --git a/openmp/libomptarget/test/mapping/target_has_device_addr.c b/openmp/libomptarget/test/mapping/target_has_device_addr.c index 897b9e4837ef..5e270e78e7d4 100644 --- a/openmp/libomptarget/test/mapping/target_has_device_addr.c +++ b/openmp/libomptarget/test/mapping/target_has_device_addr.c @@ -19,7 +19,7 @@ void foo() { A = (float *)omp_target_alloc((FROM + LENGTH) * sizeof(float), device_id); float *A_dev = NULL; -#pragma omp target has_device_addr(A [FROM:LENGTH]) map(A_dev) +#pragma omp target has_device_addr(A[FROM : LENGTH]) map(A_dev) { A_dev = A; } // CHECK: Success if (A_dev == NULL || A_dev != A) @@ -33,8 +33,8 @@ void bar() { short *xp = &x[0]; x[1] = 111; -#pragma omp target data map(tofrom : xp [0:2]) use_device_addr(xp [0:2]) -#pragma omp target has_device_addr(xp [0:2]) +#pragma omp target data map(tofrom : xp[0 : 2]) use_device_addr(xp[0 : 2]) +#pragma omp target has_device_addr(xp[0 : 2]) { xp[1] = 222; // CHECK: 222 @@ -81,7 +81,7 @@ void xoo() { short a[10], b[10]; a[1] = 111; b[1] = 111; -#pragma omp target data map(to : a [0:2], b [0:2]) use_device_addr(a, b) +#pragma omp target data map(to : a[0 : 2], b[0 : 2]) use_device_addr(a, b) #pragma omp target has_device_addr(a) has_device_addr(b[0]) { a[1] = 222; diff --git a/openmp/libomptarget/test/mapping/target_implicit_partial_map.c b/openmp/libomptarget/test/mapping/target_implicit_partial_map.c index 8df2a4d8e585..a8b2cc893df2 100644 --- a/openmp/libomptarget/test/mapping/target_implicit_partial_map.c +++ b/openmp/libomptarget/test/mapping/target_implicit_partial_map.c @@ -2,7 +2,6 @@ // RUN: %libomptarget-run-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <omp.h> @@ -11,7 +10,7 @@ int main() { int arr[100]; -#pragma omp target data map(alloc: arr[50:2]) // partially mapped +#pragma omp target data map(alloc : arr[50 : 2]) // partially mapped { // CHECK: arr[50] must present: 1 fprintf(stderr, "arr[50] must present: %d\n", @@ -30,7 +29,7 @@ int main() { arr[50] = 5; arr[51] = 6; } // must treat as present (dec ref count) even though full size not present - } // wouldn't delete if previous ref count dec didn't happen + } // wouldn't delete if previous ref count dec didn't happen // CHECK: arr[50] still present: 0 fprintf(stderr, "arr[50] still present: %d\n", diff --git a/openmp/libomptarget/test/mapping/target_pointers_members_map.cpp b/openmp/libomptarget/test/mapping/target_pointers_members_map.cpp index 6de639a0e16e..daf98199e2ca 100644 --- a/openmp/libomptarget/test/mapping/target_pointers_members_map.cpp +++ b/openmp/libomptarget/test/mapping/target_pointers_members_map.cpp @@ -38,10 +38,8 @@ int main() { // CHECK: 10 111 printf("%d %ld %p %p %p %p\n", dvb1.c[0].b.a[0], dvb1.c[0].b.d1, &dvb1, &dvb1.c[0], &dvb1.c[0].b, &dvb1.c[0].b.a[0]); -#pragma omp target map(to \ - : dvb1, dvb1.c [0:2]) \ - map(tofrom \ - : dvb1.c[0].b.a [0:10], dvb1.c[1].b.a [0:10]) +#pragma omp target map(to : dvb1, dvb1.c[0 : 2]) \ + map(tofrom : dvb1.c[0].b.a[0 : 10], dvb1.c[1].b.a[0 : 10]) { // CHECK: 10 111 printf("%d %ld %p %p %p %p\n", dvb1.c[0].b.a[0], dvb1.c[0].b.d1, &dvb1, diff --git a/openmp/libomptarget/test/mapping/target_update_array_extension.c b/openmp/libomptarget/test/mapping/target_update_array_extension.c index 9222275f27ed..c6c42bbdde1f 100644 --- a/openmp/libomptarget/test/mapping/target_update_array_extension.c +++ b/openmp/libomptarget/test/mapping/target_update_array_extension.c @@ -34,29 +34,28 @@ // RUN: %libomptarget-run-generic 2>&1 \ // RUN: | %fcheck-generic - // END. #include <stdio.h> #define BEFORE 0 -#define AFTER 1 +#define AFTER 1 #if EXTENDS == BEFORE -# define SMALL 2:3 -# define LARGE 0:5 +#define SMALL 2 : 3 +#define LARGE 0 : 5 #elif EXTENDS == AFTER -# define SMALL 0:3 -# define LARGE 0:5 +#define SMALL 0 : 3 +#define LARGE 0 : 5 #else -# error EXTENDS undefined +#error EXTENDS undefined #endif int main() { int arr[5]; // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[LARGE]) +#pragma omp target data map(alloc : arr[LARGE]) { #pragma omp target update CLAUSE(arr[SMALL]) } @@ -65,7 +64,7 @@ int main() { fprintf(stderr, "success\n"); // CHECK-NOT: Libomptarget -#pragma omp target data map(alloc: arr[SMALL]) +#pragma omp target data map(alloc : arr[SMALL]) { #pragma omp target update CLAUSE(arr[LARGE]) } diff --git a/openmp/libomptarget/test/mapping/target_use_device_addr.c b/openmp/libomptarget/test/mapping/target_use_device_addr.c index 05a5aea88a89..5c2bb8a48f6e 100644 --- a/openmp/libomptarget/test/mapping/target_use_device_addr.c +++ b/openmp/libomptarget/test/mapping/target_use_device_addr.c @@ -10,7 +10,7 @@ int main() { x[1] = 111; printf("%d, %p\n", xp[1], &xp[1]); -#pragma omp target data use_device_addr(xp [1:3]) map(tofrom : x) +#pragma omp target data use_device_addr(xp[1 : 3]) map(tofrom : x) #pragma omp target is_device_ptr(xp) { xp[1] = 222; } // CHECK: 222 diff --git a/openmp/libomptarget/test/offloading/atomic-compare-signedness.c b/openmp/libomptarget/test/offloading/atomic-compare-signedness.c index c171a2e6124d..08b7acc20338 100644 --- a/openmp/libomptarget/test/offloading/atomic-compare-signedness.c +++ b/openmp/libomptarget/test/offloading/atomic-compare-signedness.c @@ -18,11 +18,13 @@ int main() { // CHECK-NEXT: signed: xs=[[#NUM_THREADS-1]]{{$}} int xs = -1; int numThreads; - #pragma omp target parallel for num_threads(NUM_THREADS_TRY) \ - map(tofrom:xs, numThreads) +#pragma omp target parallel for num_threads(NUM_THREADS_TRY) \ + map(tofrom : xs, numThreads) for (int i = 0; i < omp_get_num_threads(); ++i) { - #pragma omp atomic compare - if (xs < i) { xs = i; } +#pragma omp atomic compare + if (xs < i) { + xs = i; + } if (i == 0) numThreads = omp_get_num_threads(); } @@ -31,11 +33,12 @@ int main() { // CHECK-NEXT: unsigned: xu=0x0{{$}} unsigned xu = UINT_MAX; - #pragma omp target parallel for num_threads(NUM_THREADS_TRY) \ - map(tofrom:xu) +#pragma omp target parallel for num_threads(NUM_THREADS_TRY) map(tofrom : xu) for (int i = 0; i < omp_get_num_threads(); ++i) { - #pragma omp atomic compare - if (xu > i) { xu = i; } +#pragma omp atomic compare + if (xu > i) { + xu = i; + } } printf("unsigned: xu=0x%x\n", xu); return 0; diff --git a/openmp/libomptarget/test/offloading/bug49021.cpp b/openmp/libomptarget/test/offloading/bug49021.cpp index feeb48e73e6c..400bbf88d735 100644 --- a/openmp/libomptarget/test/offloading/bug49021.cpp +++ b/openmp/libomptarget/test/offloading/bug49021.cpp @@ -30,8 +30,8 @@ template <typename T> int test_reduction() { sum_host += array[i]; } -#pragma omp target teams distribute parallel for map(to: array[:size]) \ - reduction(+ : sum) +#pragma omp target teams distribute parallel for map(to : array[ : size]) \ + reduction(+ : sum) for (int i = 0; i < size; i++) sum += array[i]; @@ -41,10 +41,8 @@ template <typename T> int test_reduction() { std::cout << "hierarchical parallelism" << std::endl; const int nblock(10), block_size(10); T block_sum[nblock]; -#pragma omp target teams distribute map(to \ - : array[:size]) \ - map(from \ - : block_sum[:nblock]) +#pragma omp target teams distribute map(to : array[ : size]) \ + map(from : block_sum[ : nblock]) for (int ib = 0; ib < nblock; ib++) { T partial_sum = 0; const int istart = ib * block_size; diff --git a/openmp/libomptarget/test/offloading/bug51781.c b/openmp/libomptarget/test/offloading/bug51781.c index d4a657db1eb9..17431ab6a8d8 100644 --- a/openmp/libomptarget/test/offloading/bug51781.c +++ b/openmp/libomptarget/test/offloading/bug51781.c @@ -37,18 +37,18 @@ // UNSUPPORTED: amdgcn-amd-amdhsa-LTO #if ADD_REDUCTION -# define REDUCTION(...) reduction(__VA_ARGS__) +#define REDUCTION(...) reduction(__VA_ARGS__) #else -# define REDUCTION(...) +#define REDUCTION(...) #endif #include <stdio.h> int main() { int x = 0, y = 1; - #pragma omp target teams num_teams(1) map(tofrom:x, y) REDUCTION(+:x) +#pragma omp target teams num_teams(1) map(tofrom : x, y) REDUCTION(+ : x) { x += 5; - #pragma omp parallel +#pragma omp parallel y = 6; } // CHECK: 5, 6 diff --git a/openmp/libomptarget/test/offloading/bug53727.cpp b/openmp/libomptarget/test/offloading/bug53727.cpp index 4744024dfec4..1008806fc147 100644 --- a/openmp/libomptarget/test/offloading/bug53727.cpp +++ b/openmp/libomptarget/test/offloading/bug53727.cpp @@ -22,7 +22,7 @@ int main(int argc, char *argv[]) { s.t.p[i] = i; } -#pragma omp target enter data map(to : s, s.t.p[:N]) +#pragma omp target enter data map(to : s, s.t.p[ : N]) #pragma omp target { @@ -31,14 +31,14 @@ int main(int argc, char *argv[]) { } } -#pragma omp target update from(s.t.p[:N]) +#pragma omp target update from(s.t.p[ : N]) for (int i = 0; i < N; ++i) { assert(s.t.p[i] == 2 * i); s.t.p[i] += i; } -#pragma omp target update to(s.t.p[:N]) +#pragma omp target update to(s.t.p[ : N]) #pragma omp target { @@ -47,7 +47,7 @@ int main(int argc, char *argv[]) { } } -#pragma omp target exit data map(from : s, s.t.p[:N]) +#pragma omp target exit data map(from : s, s.t.p[ : N]) for (int i = 0; i < N; ++i) { assert(s.t.p[i] == 4 * i); diff --git a/openmp/libomptarget/test/offloading/cuda_no_devices.c b/openmp/libomptarget/test/offloading/cuda_no_devices.c index ff3e3a6f5560..bbcbf6b9f07e 100644 --- a/openmp/libomptarget/test/offloading/cuda_no_devices.c +++ b/openmp/libomptarget/test/offloading/cuda_no_devices.c @@ -13,7 +13,7 @@ // CHECK-NOT: {{.}} int main() { int x = 0; - #pragma omp target teams num_teams(2) reduction(+:x) +#pragma omp target teams num_teams(2) reduction(+ : x) x += 2; printf("Hello World: %d\n", x); return 0; diff --git a/openmp/libomptarget/test/offloading/d2d_memcpy.c b/openmp/libomptarget/test/offloading/d2d_memcpy.c index 8e5a40c98dfe..7610e82291ab 100644 --- a/openmp/libomptarget/test/offloading/d2d_memcpy.c +++ b/openmp/libomptarget/test/offloading/d2d_memcpy.c @@ -27,8 +27,8 @@ int main(int argc, char *argv[]) { assert(src_ptr && "src_ptr is NULL"); assert(dst_ptr && "dst_ptr is NULL"); -#pragma omp target teams distribute parallel for device(src_device) \ - is_device_ptr(src_ptr) +#pragma omp target teams distribute parallel for device(src_device) \ + is_device_ptr(src_ptr) for (int i = 0; i < N; ++i) { src_ptr[i] = magic_num; } @@ -42,8 +42,8 @@ int main(int argc, char *argv[]) { assert(buffer && "failed to allocate host buffer"); -#pragma omp target teams distribute parallel for device(dst_device) \ - map(from: buffer[0:N]) is_device_ptr(dst_ptr) +#pragma omp target teams distribute parallel for device(dst_device) \ + map(from : buffer[0 : N]) is_device_ptr(dst_ptr) for (int i = 0; i < N; ++i) { buffer[i] = dst_ptr[i] + magic_num; } diff --git a/openmp/libomptarget/test/offloading/dynamic_module_load.c b/openmp/libomptarget/test/offloading/dynamic_module_load.c index 243598dbfd8a..935d402ef2be 100644 --- a/openmp/libomptarget/test/offloading/dynamic_module_load.c +++ b/openmp/libomptarget/test/offloading/dynamic_module_load.c @@ -19,7 +19,7 @@ int main(int argc, char **argv) { printf("dlopen() failed: %s\n", dlerror()); return 1; } - Foo = (int (*)(void)) dlsym(Handle, "foo"); + Foo = (int (*)(void))dlsym(Handle, "foo"); if (Handle == NULL) { printf("dlsym() failed: %s\n", dlerror()); return 1; diff --git a/openmp/libomptarget/test/offloading/host_as_target.c b/openmp/libomptarget/test/offloading/host_as_target.c index 963ed3888e9d..2cf2a4a9fcd4 100644 --- a/openmp/libomptarget/test/offloading/host_as_target.c +++ b/openmp/libomptarget/test/offloading/host_as_target.c @@ -11,12 +11,12 @@ // XFAIL: amdgcn-amd-amdhsa // XFAIL: amdgcn-amd-amdhsa-LTO -#include <stdio.h> #include <omp.h> +#include <stdio.h> static void check(char *X, int Dev) { printf(" host X = %c\n", *X); - #pragma omp target device(Dev) +#pragma omp target device(Dev) printf("device X = %c\n", *X); } @@ -33,62 +33,62 @@ int main(void) { // CHECK: host X = h // CHECK-NEXT: device X = d char X = 'd'; - #pragma omp target enter data map(to:X) +#pragma omp target enter data map(to : X) X = 'h'; CHECK_DATA(); - //-------------------------------------------------- - // Check behavior when specifying host directly. - //-------------------------------------------------- +//-------------------------------------------------- +// Check behavior when specifying host directly. +//-------------------------------------------------- - // CHECK-NEXT: omp_is_initial_device() = 1 - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target device(DevInit) map(always,tofrom:X) +// CHECK-NEXT: omp_is_initial_device() = 1 +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target device(DevInit) map(always, tofrom : X) printf("omp_is_initial_device() = %d\n", omp_is_initial_device()); CHECK_DATA(); - // CHECK-NEXT: omp_is_initial_device() = 1 - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target teams device(DevInit) num_teams(1) map(always,tofrom:X) +// CHECK-NEXT: omp_is_initial_device() = 1 +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target teams device(DevInit) num_teams(1) map(always, tofrom : X) printf("omp_is_initial_device() = %d\n", omp_is_initial_device()); CHECK_DATA(); - // Check that __kmpc_push_target_tripcount_mapper doesn't fail. I'm not sure - // how to check that it actually pushes to the initial device. - #pragma omp target teams device(DevInit) num_teams(1) - #pragma omp distribute +// Check that __kmpc_push_target_tripcount_mapper doesn't fail. I'm not sure +// how to check that it actually pushes to the initial device. +#pragma omp target teams device(DevInit) num_teams(1) +#pragma omp distribute for (int i = 0; i < 2; ++i) - ; + ; - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target data device(DevInit) map(always,tofrom:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target data device(DevInit) map(always, tofrom : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target enter data device(DevInit) map(always,to:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target enter data device(DevInit) map(always, to : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target exit data device(DevInit) map(always,from:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target exit data device(DevInit) map(always, from : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target update device(DevInit) to(X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target update device(DevInit) to(X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target update device(DevInit) from(X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target update device(DevInit) from(X) ; CHECK_DATA(); @@ -98,54 +98,54 @@ int main(void) { omp_set_default_device(DevInit); - // CHECK-NEXT: omp_is_initial_device() = 1 - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target map(always,tofrom:X) +// CHECK-NEXT: omp_is_initial_device() = 1 +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target map(always, tofrom : X) printf("omp_is_initial_device() = %d\n", omp_is_initial_device()); CHECK_DATA(); - // CHECK-NEXT: omp_is_initial_device() = 1 - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target teams num_teams(1) map(always,tofrom:X) +// CHECK-NEXT: omp_is_initial_device() = 1 +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target teams num_teams(1) map(always, tofrom : X) printf("omp_is_initial_device() = %d\n", omp_is_initial_device()); CHECK_DATA(); - // Check that __kmpc_push_target_tripcount_mapper doesn't fail. I'm not sure - // how to check that it actually pushes to the initial device. - #pragma omp target teams num_teams(1) - #pragma omp distribute +// Check that __kmpc_push_target_tripcount_mapper doesn't fail. I'm not sure +// how to check that it actually pushes to the initial device. +#pragma omp target teams num_teams(1) +#pragma omp distribute for (int i = 0; i < 2; ++i) - ; + ; - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target data map(always,tofrom:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target data map(always, tofrom : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target enter data map(always,to:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target enter data map(always, to : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target exit data map(always,from:X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target exit data map(always, from : X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target update to(X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target update to(X) ; CHECK_DATA(); - // CHECK-NEXT: host X = h - // CHECK-NEXT: device X = d - #pragma omp target update from(X) +// CHECK-NEXT: host X = h +// CHECK-NEXT: device X = d +#pragma omp target update from(X) ; CHECK_DATA(); diff --git a/openmp/libomptarget/test/offloading/info.c b/openmp/libomptarget/test/offloading/info.c index 006b184e1c2c..da830314964a 100644 --- a/openmp/libomptarget/test/offloading/info.c +++ b/openmp/libomptarget/test/offloading/info.c @@ -4,8 +4,8 @@ // RUN: %fcheck-nvptx64-nvidia-cuda -allow-empty -check-prefix=INFO // REQUIRES: nvptx64-nvidia-cuda -#include <stdio.h> #include <omp.h> +#include <stdio.h> #define N 64 @@ -55,14 +55,15 @@ int main() { // INFO: Libomptarget device 0 info: OpenMP Host-Device pointer mappings after block at info.c:[[#%u,]]:[[#%u,]]: // INFO: Libomptarget device 0 info: Host Ptr Target Ptr Size (B) DynRefCount HoldRefCount Declaration // INFO: Libomptarget device 0 info: [[#%#x,]] [[#%#x,]] 4 INF 0 global at unknown:0:0 -#pragma omp target data map(alloc:A[0:N]) map(ompx_hold,tofrom:B[0:N]) map(to:C[0:N]) +#pragma omp target data map(alloc : A[0 : N]) \ + map(ompx_hold, tofrom : B[0 : N]) map(to : C[0 : N]) #pragma omp target firstprivate(val) { val = 1; } __tgt_set_info_flag(0x0); // INFO-NOT: Libomptarget device 0 info: {{.*}} #pragma omp target - { } + {} return 0; } diff --git a/openmp/libomptarget/test/offloading/lone_target_exit_data.c b/openmp/libomptarget/test/offloading/lone_target_exit_data.c index 995b43d7cef5..73a5ffa2a1c2 100644 --- a/openmp/libomptarget/test/offloading/lone_target_exit_data.c +++ b/openmp/libomptarget/test/offloading/lone_target_exit_data.c @@ -8,7 +8,7 @@ int main() { // CHECK: x = 98 int x = 98; - #pragma omp target exit data map(from:x) +#pragma omp target exit data map(from : x) printf("x = %d\n", x); return 0; } diff --git a/openmp/libomptarget/test/offloading/looptripcnt.c b/openmp/libomptarget/test/offloading/looptripcnt.c index be51e12daac0..d1a095038b75 100644 --- a/openmp/libomptarget/test/offloading/looptripcnt.c +++ b/openmp/libomptarget/test/offloading/looptripcnt.c @@ -4,10 +4,9 @@ /* Test for looptripcount being popped from runtime stack. */ -#include <stdio.h> #include <omp.h> -int main() -{ +#include <stdio.h> +int main() { int N = 128; int NN = 1024; int num_teams[NN]; @@ -15,19 +14,19 @@ int main() printf("#pragma omp target teams distribute parallel for thread_limit(4)\n"); #pragma omp target teams distribute parallel for thread_limit(4) - for (int j = 0; j< N; j++) { + for (int j = 0; j < N; j++) { num_threads[j] = omp_get_num_threads(); num_teams[j] = omp_get_num_teams(); } printf("num_threads %d num_teams %d\n", num_threads[0], num_teams[0]); -// DEBUG: loop trip count is 128 + // DEBUG: loop trip count is 128 printf("#pragma omp target teams distribute parallel for\n"); #pragma omp target teams distribute parallel for - for (int j = 0; j< N; j++) { + for (int j = 0; j < N; j++) { num_threads[j] = omp_get_num_threads(); num_teams[j] = omp_get_num_teams(); } printf("num_threads %d num_teams %d\n", num_threads[0], num_teams[0]); -// DEBUG: loop trip count is 128 + // DEBUG: loop trip count is 128 return 0; } diff --git a/openmp/libomptarget/test/offloading/mandatory_but_no_devices.c b/openmp/libomptarget/test/offloading/mandatory_but_no_devices.c index 882c93579b68..365efb6a076e 100644 --- a/openmp/libomptarget/test/offloading/mandatory_but_no_devices.c +++ b/openmp/libomptarget/test/offloading/mandatory_but_no_devices.c @@ -50,7 +50,7 @@ // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory int main(void) { int X; - #pragma omp DIR device(omp_get_initial_device()) +#pragma omp DIR device(omp_get_initial_device()) ; return 0; } diff --git a/openmp/libomptarget/test/offloading/memory_manager.cpp b/openmp/libomptarget/test/offloading/memory_manager.cpp index 6a4556de554b..49759f687242 100644 --- a/openmp/libomptarget/test/offloading/memory_manager.cpp +++ b/openmp/libomptarget/test/offloading/memory_manager.cpp @@ -27,8 +27,7 @@ int main(int argc, char *argv[]) { } int buffer[n]; #pragma omp target teams distribute parallel for is_device_ptr(p) \ - map(from \ - : buffer) + map(from : buffer) for (int j = 0; j < n; ++j) { buffer[j] = p[j]; } diff --git a/openmp/libomptarget/test/offloading/non_contiguous_update.cpp b/openmp/libomptarget/test/offloading/non_contiguous_update.cpp index 91d64493d605..609f0f967fb1 100644 --- a/openmp/libomptarget/test/offloading/non_contiguous_update.cpp +++ b/openmp/libomptarget/test/offloading/non_contiguous_update.cpp @@ -1,9 +1,9 @@ // RUN: %libomptarget-compile-generic && env LIBOMPTARGET_DEBUG=1 %libomptarget-run-generic 2>&1 | %fcheck-generic -allow-empty -check-prefix=DEBUG // REQUIRES: libomptarget-debug +#include <cassert> #include <cstdio> #include <cstdlib> -#include <cassert> // Data structure definitions copied from OpenMP RTL. struct __tgt_target_non_contig { @@ -12,9 +12,7 @@ struct __tgt_target_non_contig { int64_t stride; }; -enum tgt_map_type { - OMP_TGT_MAPTYPE_NON_CONTIG = 0x100000000000 -}; +enum tgt_map_type { OMP_TGT_MAPTYPE_NON_CONTIG = 0x100000000000 }; // OpenMP RTL interfaces #ifdef __cplusplus @@ -93,4 +91,3 @@ int main() { // DEBUG: offset 456 return 0; } - diff --git a/openmp/libomptarget/test/offloading/offloading_success.c b/openmp/libomptarget/test/offloading/offloading_success.c index 919972a160b0..f849cb4cfba7 100644 --- a/openmp/libomptarget/test/offloading/offloading_success.c +++ b/openmp/libomptarget/test/offloading/offloading_success.c @@ -1,12 +1,12 @@ // RUN: %libomptarget-compile-run-and-check-generic -#include <stdio.h> #include <omp.h> +#include <stdio.h> int main(void) { int isHost = -1; -#pragma omp target map(from: isHost) +#pragma omp target map(from : isHost) { isHost = omp_is_initial_device(); } if (isHost < 0) { diff --git a/openmp/libomptarget/test/offloading/offloading_success.cpp b/openmp/libomptarget/test/offloading/offloading_success.cpp index bc049a6a2c45..d3103b8666ab 100644 --- a/openmp/libomptarget/test/offloading/offloading_success.cpp +++ b/openmp/libomptarget/test/offloading/offloading_success.cpp @@ -1,12 +1,12 @@ // RUN: %libomptarget-compilexx-run-and-check-generic -#include <stdio.h> #include <omp.h> +#include <stdio.h> int main(void) { int isHost = 0; -#pragma omp target map(from: isHost) +#pragma omp target map(from : isHost) { isHost = omp_is_initial_device(); } if (isHost < 0) { diff --git a/openmp/libomptarget/test/offloading/requires.c b/openmp/libomptarget/test/offloading/requires.c index 25c5053699b2..e5e58012a37c 100644 --- a/openmp/libomptarget/test/offloading/requires.c +++ b/openmp/libomptarget/test/offloading/requires.c @@ -8,8 +8,8 @@ (for whatever reason) the set must be consistent with previously set values. */ -#include <stdio.h> #include <omp.h> +#include <stdio.h> // --------------------------------------------------------------------------- // Various definitions copied from OpenMP RTL diff --git a/openmp/libomptarget/test/offloading/target-teams-atomic.c b/openmp/libomptarget/test/offloading/target-teams-atomic.c index 5d482f1cc1be..97563422ced8 100644 --- a/openmp/libomptarget/test/offloading/target-teams-atomic.c +++ b/openmp/libomptarget/test/offloading/target-teams-atomic.c @@ -17,9 +17,9 @@ int main() { // CHECK-NEXT: update: x=[[#NUM_TEAMS]]{{$}} int x = 0; int numTeams; - #pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom:x, numTeams) +#pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom : x, numTeams) { - #pragma omp atomic update +#pragma omp atomic update ++x; if (omp_get_team_num() == 0) numTeams = omp_get_num_teams(); @@ -32,10 +32,10 @@ int main() { bool xCaptured[numTeams]; memset(xCaptured, 0, sizeof xCaptured); x = 0; - #pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom:x, numTeams) +#pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom : x, numTeams) { int v; - #pragma omp atomic capture +#pragma omp atomic capture v = x++; xCaptured[v] = true; } diff --git a/openmp/libomptarget/test/offloading/target_depend_nowait.cpp b/openmp/libomptarget/test/offloading/target_depend_nowait.cpp index 54b212c98921..b0d5408770bc 100644 --- a/openmp/libomptarget/test/offloading/target_depend_nowait.cpp +++ b/openmp/libomptarget/test/offloading/target_depend_nowait.cpp @@ -40,10 +40,8 @@ int main() { #pragma omp target update from(A) depend(out : A[0]) nowait // B updated via exit, A just released -#pragma omp target exit data map(release \ - : A) map(from \ - : B) depend(out \ - : A[0]) nowait +#pragma omp target exit data map(release : A) map(from : B) depend(out : A[0]) \ + nowait } // if } // parallel @@ -56,4 +54,3 @@ int main() { return Sum != 2 * N * (2 + N - 1 + 2) / 2; } - diff --git a/openmp/libomptarget/test/offloading/target_nowait_target.cpp b/openmp/libomptarget/test/offloading/target_nowait_target.cpp index 24a83c300524..c00b23d316d9 100644 --- a/openmp/libomptarget/test/offloading/target_nowait_target.cpp +++ b/openmp/libomptarget/test/offloading/target_nowait_target.cpp @@ -11,14 +11,15 @@ int main(int argc, char *argv[]) { for (int i = 0; i < 1024; ++i) data[i] = i; -#pragma omp target map(tofrom: sum) map(to: data) depend(inout : data[0]) nowait +#pragma omp target map(tofrom : sum) map(to : data) depend(inout : data[0]) \ + nowait { for (int i = 0; i < 1024; ++i) { sum += data[i]; } } -#pragma omp target map(tofrom: sum) map(to: data) depend(inout : data[0]) +#pragma omp target map(tofrom : sum) map(to : data) depend(inout : data[0]) { for (int i = 0; i < 1024; ++i) { sum += data[i]; diff --git a/openmp/libomptarget/test/offloading/test_libc.cpp b/openmp/libomptarget/test/offloading/test_libc.cpp index 66d73d7b3e2b..859c2da97de0 100644 --- a/openmp/libomptarget/test/offloading/test_libc.cpp +++ b/openmp/libomptarget/test/offloading/test_libc.cpp @@ -7,7 +7,7 @@ extern "C" int printf(const char *, ...); // std::equal is lowered to libc function memcmp. void test_memcpy() { int r = 0; -#pragma omp target map(from: r) +#pragma omp target map(from : r) { int x[2] = {0, 0}; int y[2] = {0, 0}; diff --git a/openmp/libomptarget/test/offloading/wtime.c b/openmp/libomptarget/test/offloading/wtime.c index c1cff80b7c7f..84118deb02ec 100644 --- a/openmp/libomptarget/test/offloading/wtime.c +++ b/openmp/libomptarget/test/offloading/wtime.c @@ -12,7 +12,7 @@ int main(int argc, char *argv[]) { int *data = (int *)malloc(N * sizeof(int)); -#pragma omp target map(from: data[0:N]) +#pragma omp target map(from : data[0 : N]) { double start = omp_get_wtime(); for (int i = 0; i < N; ++i) @@ -26,4 +26,3 @@ int main(int argc, char *argv[]) { } // CHECK: duration: {{.+[1-9]+}} - diff --git a/openmp/libomptarget/test/unified_shared_memory/api.c b/openmp/libomptarget/test/unified_shared_memory/api.c index 3b30843f0b44..79c9f6ebedff 100644 --- a/openmp/libomptarget/test/unified_shared_memory/api.c +++ b/openmp/libomptarget/test/unified_shared_memory/api.c @@ -6,8 +6,8 @@ // UNSUPPORTED: amdgcn-amd-amdhsa // UNSUPPORTED: amdgcn-amd-amdhsa-LTO -#include <stdio.h> #include <omp.h> +#include <stdio.h> // --------------------------------------------------------------------------- // Various definitions copied from OpenMP RTL diff --git a/openmp/libomptarget/test/unified_shared_memory/associate_ptr.c b/openmp/libomptarget/test/unified_shared_memory/associate_ptr.c index 1e1b68f2d9b0..5f795dd42dbd 100644 --- a/openmp/libomptarget/test/unified_shared_memory/associate_ptr.c +++ b/openmp/libomptarget/test/unified_shared_memory/associate_ptr.c @@ -17,11 +17,11 @@ int main(int argc, char *argv[]) { int rc = omp_target_associate_ptr(&x, x_dev, sizeof x, 0, dev); assert(!rc && "expected omp_target_associate_ptr to succeed"); - // To determine whether x needs to be transfered, the runtime cannot simply - // check whether unified shared memory is enabled and the 'close' modifier is - // specified. It must check whether x was previously placed in device memory - // by, for example, omp_target_associate_ptr. - #pragma omp target map(always, tofrom: x) +// To determine whether x needs to be transfered, the runtime cannot simply +// check whether unified shared memory is enabled and the 'close' modifier is +// specified. It must check whether x was previously placed in device memory +// by, for example, omp_target_associate_ptr. +#pragma omp target map(always, tofrom : x) x += 1; // CHECK: x=11 diff --git a/openmp/libomptarget/test/unified_shared_memory/close_member.c b/openmp/libomptarget/test/unified_shared_memory/close_member.c index a41e4d8a0546..e5a15e3d19ab 100644 --- a/openmp/libomptarget/test/unified_shared_memory/close_member.c +++ b/openmp/libomptarget/test/unified_shared_memory/close_member.c @@ -17,20 +17,20 @@ int main(int argc, char *argv[]) { int dev = omp_get_default_device(); struct S s = {10, 20}; - #pragma omp target enter data map(close, to: s) - #pragma omp target map(alloc: s) +#pragma omp target enter data map(close, to : s) +#pragma omp target map(alloc : s) { s.x = 11; s.y = 21; } - // To determine whether x needs to be transfered or deleted, the runtime - // cannot simply check whether unified shared memory is enabled and the - // 'close' modifier is specified. It must check whether x was previously - // placed in device memory by, for example, a 'close' modifier that isn't - // specified here. The following struct member case checks a special code - // path in the runtime implementation where members are transferred before - // deletion of the struct. - #pragma omp target exit data map(from: s.x, s.y) +// To determine whether x needs to be transfered or deleted, the runtime +// cannot simply check whether unified shared memory is enabled and the +// 'close' modifier is specified. It must check whether x was previously +// placed in device memory by, for example, a 'close' modifier that isn't +// specified here. The following struct member case checks a special code +// path in the runtime implementation where members are transferred before +// deletion of the struct. +#pragma omp target exit data map(from : s.x, s.y) // CHECK: s.x=11, s.y=21 printf("s.x=%d, s.y=%d\n", s.x, s.y); diff --git a/openmp/libomptarget/test/unified_shared_memory/close_modifier.c b/openmp/libomptarget/test/unified_shared_memory/close_modifier.c index ce368a359176..47902a9a298d 100644 --- a/openmp/libomptarget/test/unified_shared_memory/close_modifier.c +++ b/openmp/libomptarget/test/unified_shared_memory/close_modifier.c @@ -6,7 +6,6 @@ // amdgpu runtime crash // UNSUPPORTED: amdgcn-amd-amdhsa - #include <omp.h> #include <stdio.h> @@ -33,10 +32,8 @@ int main(int argc, char *argv[]) { // Test that updates on the device are not visible to host // when only a TO mapping is used. // -#pragma omp target map(tofrom \ - : device_data, device_alloc) map(close, to \ - : alloc[:N], data \ - [:N]) +#pragma omp target map(tofrom : device_data, device_alloc) \ + map(close, to : alloc[ : N], data[ : N]) { device_data = &data[0]; device_alloc = &alloc[0]; @@ -84,7 +81,7 @@ int main(int argc, char *argv[]) { data[i] += 1; } -#pragma omp target map(close, tofrom : alloc[:N], data[:N]) +#pragma omp target map(close, tofrom : alloc[ : N], data[ : N]) { // CHECK: Alloc device values are correct: Succeeded fails = 0; diff --git a/openmp/libomptarget/test/unified_shared_memory/shared_update.c b/openmp/libomptarget/test/unified_shared_memory/shared_update.c index b211d333453e..3923b89bba4f 100644 --- a/openmp/libomptarget/test/unified_shared_memory/shared_update.c +++ b/openmp/libomptarget/test/unified_shared_memory/shared_update.c @@ -5,8 +5,8 @@ // amdgpu runtime crash // UNSUPPORTED: amdgcn-amd-amdhsa -#include <stdio.h> #include <omp.h> +#include <stdio.h> // --------------------------------------------------------------------------- // Various definitions copied from OpenMP RTL |