This patch implements Clang front end support for the OpenMP TR8 `present` map type modifier. The next patch in this series implements OpenMP runtime support. This patch does not attempt to implement TR8 sec. 2.22.7.1 "map Clause", p. 319, L14-16: > If a map clause with a present map-type-modifier is present in a map > clause, then the effect of the clause is ordered before all other > map clauses that do not have the present modifier. Compare to L10-11, which Clang does not appear to implement yet: > For a given construct, the effect of a map clause with the to, from, > or tofrom map-type is ordered before the effect of a map clause with > the alloc, release, or delete map-type. This patch also does not implement the `present` implicit-behavior for `defaultmap` or the `present` motion-modifier for `target update`. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D83061
246 lines
8.0 KiB
C++
246 lines
8.0 KiB
C++
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -ast-print %s | FileCheck %s
|
|
// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
|
|
// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
|
|
|
|
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s | FileCheck %s
|
|
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
|
|
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
|
|
|
|
// RUN: %clang_cc1 -DOMP51 -verify -fopenmp -fopenmp-version=51 -ast-print %s | FileCheck -check-prefixes=CHECK,OMP51 %s
|
|
// RUN: %clang_cc1 -DOMP51 -fopenmp -fopenmp-version=51 -x c++ -std=c++11 -emit-pch -o %t %s
|
|
// RUN: %clang_cc1 -DOMP51 -fopenmp -fopenmp-version=51 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck -check-prefixes=CHECK,OMP51 %s
|
|
|
|
// RUN: %clang_cc1 -DOMP51 -verify -fopenmp-simd -fopenmp-version=51 -ast-print %s | FileCheck -check-prefixes=CHECK,OMP51 %s
|
|
// RUN: %clang_cc1 -DOMP51 -fopenmp-simd -fopenmp-version=51 -x c++ -std=c++11 -emit-pch -o %t %s
|
|
// RUN: %clang_cc1 -DOMP51 -fopenmp-simd -fopenmp-version=51 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck -check-prefixes=CHECK,OMP51 %s
|
|
// expected-no-diagnostics
|
|
|
|
#ifndef HEADER
|
|
#define HEADER
|
|
|
|
void foo() {}
|
|
|
|
template <typename T, int C>
|
|
T tmain(T argc, T *argv) {
|
|
T i, j, b, c, d, e, x[20];
|
|
|
|
#pragma omp target data map(to: c)
|
|
i = argc;
|
|
|
|
#pragma omp target data map(to: c) if (target data: j > 0)
|
|
foo();
|
|
|
|
#pragma omp target data map(to: c) if (b)
|
|
foo();
|
|
|
|
#pragma omp target data map(c)
|
|
foo();
|
|
|
|
#pragma omp target data map(c) if(b>e)
|
|
foo();
|
|
|
|
#pragma omp target data map(x[0:10], c)
|
|
foo();
|
|
|
|
#pragma omp target data map(to: c) map(from: d)
|
|
foo();
|
|
|
|
#pragma omp target data map(always,alloc: e)
|
|
foo();
|
|
|
|
#pragma omp target data map(close,alloc: e)
|
|
foo();
|
|
|
|
#ifdef OMP51
|
|
#pragma omp target data map(present,alloc: e)
|
|
foo();
|
|
#endif
|
|
|
|
// nesting a target region
|
|
#pragma omp target data map(e)
|
|
{
|
|
#pragma omp target map(always, alloc: e)
|
|
foo();
|
|
#pragma omp target map(close, alloc: e)
|
|
foo();
|
|
#ifdef OMP51
|
|
#pragma omp target map(present, alloc: e)
|
|
foo();
|
|
#endif
|
|
}
|
|
|
|
return 0;
|
|
}
|
|
|
|
// CHECK: template <typename T, int C> T tmain(T argc, T *argv) {
|
|
// CHECK-NEXT: T i, j, b, c, d, e, x[20];
|
|
// CHECK-NEXT: #pragma omp target data map(to: c){{$}}
|
|
// CHECK-NEXT: i = argc;
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(target data: j > 0)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(b)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c) if(b > e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: x[0:10],c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) map(from: d)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target data map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: e)
|
|
// CHECK-NEXT: {
|
|
// CHECK-NEXT: #pragma omp target map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
// CHECK: template<> int tmain<int, 5>(int argc, int *argv) {
|
|
// CHECK-NEXT: int i, j, b, c, d, e, x[20];
|
|
// CHECK-NEXT: #pragma omp target data map(to: c)
|
|
// CHECK-NEXT: i = argc;
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(target data: j > 0)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(b)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c) if(b > e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: x[0:10],c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) map(from: d)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target data map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: e)
|
|
// CHECK-NEXT: {
|
|
// CHECK-NEXT: #pragma omp target map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
// CHECK: template<> char tmain<char, 1>(char argc, char *argv) {
|
|
// CHECK-NEXT: char i, j, b, c, d, e, x[20];
|
|
// CHECK-NEXT: #pragma omp target data map(to: c)
|
|
// CHECK-NEXT: i = argc;
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(target data: j > 0)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) if(b)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c) if(b > e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: x[0:10],c)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) map(from: d)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target data map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: e)
|
|
// CHECK-NEXT: {
|
|
// CHECK-NEXT: #pragma omp target map(always,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// CHECK-NEXT: #pragma omp target map(close,alloc: e)
|
|
// CHECK-NEXT: foo();
|
|
// OMP51-NEXT: #pragma omp target map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
|
|
int main (int argc, char **argv) {
|
|
int b = argc, c, d, e, f, g, x[20];
|
|
static int a;
|
|
// CHECK: static int a;
|
|
|
|
#pragma omp target data map(to: ([argc][3][a])argv)
|
|
// CHECK: #pragma omp target data map(to: ([argc][3][a])argv)
|
|
#pragma omp target data map(to: c)
|
|
// CHECK: #pragma omp target data map(to: c)
|
|
a=2;
|
|
// CHECK-NEXT: a = 2;
|
|
#pragma omp target data map(to: c) if (target data: b)
|
|
// CHECK: #pragma omp target data map(to: c) if(target data: b)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(to: c) if (b > g)
|
|
// CHECK: #pragma omp target data map(to: c) if(b > g)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(c)
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(c) if(b>g)
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: c) if(b > g)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(x[0:10], c)
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: x[0:10],c)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(to: c) map(from: d)
|
|
// CHECK-NEXT: #pragma omp target data map(to: c) map(from: d)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(always,alloc: e)
|
|
// CHECK-NEXT: #pragma omp target data map(always,alloc: e)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
#pragma omp target data map(close,alloc: e)
|
|
// CHECK-NEXT: #pragma omp target data map(close,alloc: e)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
|
|
// OMP51-NEXT: #pragma omp target data map(present,alloc: e)
|
|
// OMP51-NEXT: foo();
|
|
#ifdef OMP51
|
|
#pragma omp target data map(present,alloc: e)
|
|
foo();
|
|
#endif
|
|
|
|
// nesting a target region
|
|
#pragma omp target data map(e)
|
|
// CHECK-NEXT: #pragma omp target data map(tofrom: e)
|
|
{
|
|
// CHECK-NEXT: {
|
|
#pragma omp target map(always, alloc: e)
|
|
// CHECK-NEXT: #pragma omp target map(always,alloc: e)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
#pragma omp target map(close, alloc: e)
|
|
// CHECK-NEXT: #pragma omp target map(close,alloc: e)
|
|
foo();
|
|
// CHECK-NEXT: foo();
|
|
#pragma omp target map(always, alloc: e)
|
|
// CHECK-NEXT: #pragma omp target map(always,alloc: e)
|
|
foo();
|
|
}
|
|
|
|
return tmain<int, 5>(argc, &argc) + tmain<char, 1>(argv[0][0], argv[0]);
|
|
}
|
|
|
|
#endif
|