cchen updated this revision to Diff 308760.
cchen added a comment.
Rebase and fix coding style
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D92427/new/
https://reviews.llvm.org/D92427
Files:
clang/include/clang/Basic/OpenMPKinds.def
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/target_ast_print.cpp
clang/test/OpenMP/target_defaultmap_codegen.cpp
clang/test/OpenMP/target_defaultmap_messages.cpp
Index: clang/test/OpenMP/target_defaultmap_messages.cpp
===================================================================
--- clang/test/OpenMP/target_defaultmap_messages.cpp
+++ clang/test/OpenMP/target_defaultmap_messages.cpp
@@ -1,5 +1,8 @@
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=51 %s -verify=expected,omp51 -Wuninitialized -DOMP51
+// run: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=51 %s -verify=expected,omp51 -wuninitialized -domp51
+
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -verify=expected,omp5 -Wuninitialized -DOMP5
-// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -verify=expected,omp5 -Wuninitialized -DOMP5
+// run: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -verify=expected,omp5 -wuninitialized -domp5
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 %s -verify=expected,omp45 -Wuninitialized
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 %s -verify=expected,omp45 -Wuninitialized
@@ -27,23 +30,24 @@
T tmain(T argc, S **argv) {
#pragma omp target defaultmap // expected-error {{expected '(' after 'defaultmap'}}
foo();
-#pragma omp target defaultmap( // omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap( // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
-#pragma omp target defaultmap() // omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap() // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom // expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
- #pragma omp target defaultmap (tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
+ #pragma omp target defaultmap (tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp51-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom) // omp45-warning {{missing ':' after defaultmap modifier - ignoring}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom, // expected-error {{expected ')'}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
- #pragma omp target defaultmap (scalar: // omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+ #pragma omp target defaultmap (scalar: // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp51-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom, scalar // expected-error {{expected ')'}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
- #pragma omp target defaultmap(tofrom:scalar) defaultmap(tofrom:scalar) // omp45-error {{directive '#pragma omp target' cannot contain more than one 'defaultmap' clause}} omp5-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
+ #pragma omp target defaultmap(tofrom:scalar) defaultmap(tofrom:scalar) // omp45-error {{directive '#pragma omp target' cannot contain more than one 'defaultmap' clause}} omp5-error {{at most one defaultmap clause for each variable-category can appear on the directive}} omp51-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
+
foo();
#if OMP5
@@ -76,29 +80,36 @@
baz.bar.a += argc; // expected-error {{variable 'baz' must have explicitly specified data sharing attributes, data mapping attributes, or in an is_device_ptr clause}}
#endif
+#if OMP51
+#pragma omp target defaultmap(present: somthing) // omp51-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}}}
+ foo();
+#pragma omp target defaultmap(present: scalar) defaultmap(present: scalar) // omp51-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
+ foo();
+#endif
+
return argc;
}
int main(int argc, char **argv) {
#pragma omp target defaultmap // expected-error {{expected '(' after 'defaultmap'}}
foo();
-#pragma omp target defaultmap( // omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap( // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
-#pragma omp target defaultmap() // omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap() // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom // expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
-#pragma omp target defaultmap(tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap(tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp51-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom) // omp45-warning {{missing ':' after defaultmap modifier - ignoring}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom, // expected-error {{expected ')'}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
-#pragma omp target defaultmap(scalar: // omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
+#pragma omp target defaultmap(scalar: // omp51-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp51-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} omp5-error {{expected 'scalar', 'aggregate', 'pointer' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} omp5-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} omp45-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}}
foo();
#pragma omp target defaultmap(tofrom, scalar // expected-error {{expected ')'}} omp45-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} omp45-error {{expected 'scalar' in OpenMP clause 'defaultmap'}}
foo();
-#pragma omp target defaultmap(tofrom: scalar) defaultmap(tofrom: scalar) // omp45-error {{directive '#pragma omp target' cannot contain more than one 'defaultmap' clause}} omp5-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
+#pragma omp target defaultmap(tofrom: scalar) defaultmap(tofrom: scalar) // omp45-error {{directive '#pragma omp target' cannot contain more than one 'defaultmap' clause}} omp5-error {{at most one defaultmap clause for each variable-category can appear on the directive}} omp51-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
foo();
#ifdef OMP5
Index: clang/test/OpenMP/target_defaultmap_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_defaultmap_codegen.cpp
+++ clang/test/OpenMP/target_defaultmap_codegen.cpp
@@ -1520,7 +1520,7 @@
// SIMD-ONLY18-NOT: {{__kmpc|__tgt}}
#ifdef CK26
-// CK26-DAG: [[SIZES:@.+]] = {{.+}}constant [3 x i64] [i64 4096, i64 4, i64 {{.+}}]
+// CK26-DAG: [[SIZES:@.+]] = {{.+}}constant [3 x i64] [i64 4, i64 4096, i64 {{.+}}]
// Map types: OMP_MAP_TO | MAP_ALWAYS | OMP_MAP_IMPLICIT = 533
// CK26-DAG: [[TYPES:@.+]] = {{.+}}constant [3 x i64] [i64 531, i64 531, i64 531]
@@ -1553,5 +1553,191 @@
}
}
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK27 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK27 --check-prefix CK27-64
+// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK27 --check-prefix CK27-64
+// RUN: %clang_cc1 -DCK27 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK27 --check-prefix CK27-32
+// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK27 --check-prefix CK27-32
+
+// RUN: %clang_cc1 -DCK27 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK27 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK27 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK27 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// SIMD-ONLY6-NOT: {{__kmpc|__tgt}}
+#ifdef CK27
+
+// CK27-LABEL: @.__omp_offloading_{{.*}}implicit_present_scalar{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
+
+// CK27-DAG: [[SIZES:@.+]] = {{.+}}constant [1 x i64] [i64 4]
+// Map types: OMP_MAP_TARGET_PARAM | OMP_MAP_IMPLICIT | OMP_MAP_PRESENT = 4640
+// CK27-DAG: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 4640]
+
+// CK27-LABEL: implicit_present_scalar{{.*}}(
+void implicit_present_scalar(int a) {
+ // CK27-DAG: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}}, i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}}, i8** null, i8** null)
+ // CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0
+ // CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0
+ // CK27-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0
+ // CK27-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0
+ // CK27-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to i32**
+ // CK27-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
+ // CK27-DAG: store i32* [[DECL:%[^,]+]], i32** [[CBP1]]
+ // CK27-DAG: store i32* [[DECL]], i32** [[CP1]]
+ #pragma omp target defaultmap(present: scalar)
+ {
+ a += 1.0;
+ }
+}
+
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK28 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK28 --check-prefix CK28-64
+// RUN: %clang_cc1 -DCK28 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK28 --check-prefix CK28-64
+// RUN: %clang_cc1 -DCK28 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK28 --check-prefix CK28-32
+// RUN: %clang_cc1 -DCK28 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK28 --check-prefix CK28-32
+
+// RUN: %clang_cc1 -DCK28 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK28 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK28 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// RUN: %clang_cc1 -DCK28 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY6 %s
+// SIMD-ONLY6-NOT: {{__kmpc|__tgt}}
+#ifdef CK28
+
+// CK28-LABEL: @.__omp_offloading_{{.*}}implicit_present_aggregate{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
+
+// CK28-DAG: [[SIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
+// Map types: OMP_MAP_TARGET_PARAM | OMP_MAP_IMPLICIT | OMP_MAP_PRESENT = 4640
+// CK28-DAG: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 4640]
+
+// CK28-LABEL: implicit_present_aggregate{{.*}}(
+void implicit_present_aggregate(int a) {
+ double darr[2] = {(double)a, (double)a};
+
+ // CK28-DAG: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}}, i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}}, i8** null, i8** null)
+ // CK28-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0
+ // CK28-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0
+ // CK28-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0
+ // CK28-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0
+ // CK28-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [2 x double]**
+ // CK28-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [2 x double]**
+ // CK28-DAG: store [2 x double]* [[DECL:%[^,]+]], [2 x double]** [[CBP1]]
+ // CK28-DAG: store [2 x double]* [[DECL]], [2 x double]** [[CP1]]
+
+ // CK28: call void [[KERNEL:@.+]]([2 x double]* [[DECL]])
+ #pragma omp target defaultmap(present: aggregate)
+ {
+ darr[0] += 1.0;
+ darr[1] += 1.0;
+ }
+}
+
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK29 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK29
+// RUN: %clang_cc1 -DCK29 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK29
+// RUN: %clang_cc1 -DCK29 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK29
+// RUN: %clang_cc1 -DCK29 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK29
+
+// RUN: %clang_cc1 -DCK29 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY8 %s
+// RUN: %clang_cc1 -DCK29 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY8 %s
+// RUN: %clang_cc1 -DCK29 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY8 %s
+// RUN: %clang_cc1 -DCK29 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY8 %s
+// SIMD-ONLY8-NOT: {{__kmpc|__tgt}}
+#ifdef CK29
+
+
+// CK29-LABEL: @.__omp_offloading_{{.*}}explicit_present_pointer{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
+
+// CK29: [[SIZE:@.+]] = private {{.*}}constant [1 x i64] [i64 {{8|4}}]
+// Map types: OMP_MAP_TARGET_PARAM | OMP_MAP_IMPLICIT | OMP_MAP_PRESENT = 4640
+// CK29: [[MTYPE:@.+]] = private {{.*}}constant [1 x i64] [i64 4640]
+
+// CK29-LABEL: explicit_present_pointer{{.*}}(
+void explicit_present_pointer() {
+ int *pa;
+
+ // CK29-DAG: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}}, i64 {{[^,]+}}, i8* {{[^,]+}}, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE]]{{.+}}, i8** null, i8** null)
+ // CK29-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK29-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+
+ // CK29-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
+ // CK29-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
+ // CK29-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32***
+ // CK29-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32***
+ // CK29-DAG: store i32** [[VAR0:%.+]], i32*** [[CBP0]]
+ // CK29-DAG: store i32** [[VAR0]], i32*** [[CP0]]
+
+ #pragma omp target defaultmap(present: pointer)
+ {
+ pa[50]++;
+ }
+}
+
+#endif
+#ifdef CK30
+
+///==========================================================================///
+// RUN: %clang_cc1 -DCK30 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK30
+// RUN: %clang_cc1 -DCK30 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK30
+// RUN: %clang_cc1 -DCK30 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK30
+// RUN: %clang_cc1 -DCK30 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap %s --check-prefix CK30
+
+// RUN: %clang_cc1 -DCK30 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY10 %s
+// RUN: %clang_cc1 -DCK30 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY10 %s
+// RUN: %clang_cc1 -DCK30 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY10 %s
+// RUN: %clang_cc1 -DCK30 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap --check-prefix SIMD-ONLY10 %s
+// SIMD-ONLY10-NOT: {{__kmpc|__tgt}}
+
+// CK30-LABEL: @.__omp_offloading_{{.*}}implicit_present_double_complex{{.*}}.region_id = weak constant i8 0
+
+// CK30-DAG: [[SIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16]
+// Map types: OMP_MAP_TARGET_PARAM | OMP_MAP_IMPLICIT | OMP_MAP_PRESENT = 4640
+// CK30-DAG: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 4640]
+
+// CK30-LABEL: implicit_present_double_complex{{.*}}(
+void implicit_present_double_complex (int a){
+ double _Complex dc = (double)a;
+
+ // CK30-DAG: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}}, i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}}, i8** null, i8** null)
+ // CK30-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0
+ // CK30-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0
+ // CK30-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0
+ // CK30-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0
+ // CK30-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to { double, double }**
+ // CK30-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to { double, double }**
+ // CK30-DAG: store { double, double }* [[PTR:%[^,]+]], { double, double }** [[CBP1]]
+ // CK30-DAG: store { double, double }* [[PTR]], { double, double }** [[CP1]]
+
+ // CK30: call void [[KERNEL:@.+]]({ double, double }* [[PTR]])
+ #pragma omp target defaultmap(present:scalar)
+ {
+ dc *= dc;
+ }
+}
+
+// CK30: define internal void [[KERNEL]]({ double, double }* {{.*}}[[ARG:%.+]])
+// CK30: [[ADDR:%.+]] = alloca { double, double }*,
+// CK30: store { double, double }* [[ARG]], { double, double }** [[ADDR]],
+// CK30: [[REF:%.+]] = load { double, double }*, { double, double }** [[ADDR]],
+// CK30: {{.+}} = getelementptr inbounds { double, double }, { double, double }* [[REF]], i32 0, i32 0
#endif
#endif
Index: clang/test/OpenMP/target_ast_print.cpp
===================================================================
--- clang/test/OpenMP/target_ast_print.cpp
+++ clang/test/OpenMP/target_ast_print.cpp
@@ -1073,4 +1073,56 @@
}
#endif //OMP5
+
+#ifdef OMP51
+
+///==========================================================================///
+// RUN: %clang_cc1 -DOMP51 -verify -fopenmp -fopenmp-version=51 -ast-print %s | FileCheck %s --check-prefix OMP51
+// 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 %s --check-prefix OMP51
+
+// RUN: %clang_cc1 -DOMP51 -verify -fopenmp-simd -fopenmp-version=51 -ast-print %s | FileCheck %s --check-prefix OMP51
+// 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 %s --check-prefix OMP51
+
+void foo() {}
+
+template <typename T, int C>
+T tmain(T argc, T *argv) {
+ #pragma omp target defaultmap(present: scalar)
+ foo();
+ #pragma omp target defaultmap(present: aggregate)
+ foo();
+ #pragma omp target defaultmap(present: pointer)
+ foo();
+
+ return 0;
+}
+
+// OMP51: template <typename T, int C> T tmain(T argc, T *argv) {
+// OMP51-NEXT: #pragma omp target defaultmap(present: scalar)
+// OMP51-NEXT: foo()
+// OMP51-NEXT: #pragma omp target defaultmap(present: aggregate)
+// OMP51-NEXT: foo()
+// OMP51-NEXT: #pragma omp target defaultmap(present: pointer)
+// OMP51-NEXT: foo()
+
+// OMP51-LABEL: int main(int argc, char **argv) {
+int main (int argc, char **argv) {
+#pragma omp target defaultmap(present: scalar)
+// OMP51-NEXT: #pragma omp target defaultmap(present: scalar)
+ foo();
+// OMP51-NEXT: foo();
+#pragma omp target defaultmap(present: aggregate)
+// OMP51-NEXT: #pragma omp target defaultmap(present: aggregate)
+ foo();
+// OMP51-NEXT: foo();
+#pragma omp target defaultmap(present: pointer)
+// OMP51-NEXT: #pragma omp target defaultmap(present: pointer)
+ foo();
+// OMP51-NEXT: foo();
+
+ return tmain<int, 5>(argc, &argc) + tmain<char, 1>(argv[0][0], argv[0]);
+}
+#endif
#endif
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -3306,6 +3306,13 @@
case OMPC_DEFAULTMAP_MODIFIER_tofrom:
Kind = OMPC_MAP_tofrom;
break;
+ case OMPC_DEFAULTMAP_MODIFIER_present:
+ // If implicit-behavior is present, each variable referenced in the
+ // construct in the category specified by variable-category is treated as if
+ // it had been listed in a map clause with the map-type of alloc and
+ // map-type-modifier of present.
+ Kind = OMPC_MAP_alloc;
+ break;
case OMPC_DEFAULTMAP_MODIFIER_firstprivate:
case OMPC_DEFAULTMAP_MODIFIER_last:
llvm_unreachable("Unexpected defaultmap implicit behavior");
@@ -3332,8 +3339,13 @@
bool ErrorFound = false;
bool TryCaptureCXXThisMembers = false;
CapturedStmt *CS = nullptr;
+ const static unsigned DefaultmapKindNum = OMPC_DEFAULTMAP_pointer + 1;
llvm::SmallVector<Expr *, 4> ImplicitFirstprivate;
- llvm::SmallVector<Expr *, 4> ImplicitMap[OMPC_MAP_delete];
+ llvm::SmallVector<Expr *, 4> ImplicitMap[DefaultmapKindNum][OMPC_MAP_delete];
+ llvm::SmallVector<OpenMPMapModifierKind, NumberOfOMPMapClauseModifiers>
+ ImplicitMapModifier[DefaultmapKindNum];
+ llvm::SmallVector<SourceLocation, NumberOfOMPMapClauseModifiers>
+ ImplicitMapModifierLoc[DefaultmapKindNum];
Sema::VarsWithInheritedDSAType VarsWithInheritedDSA;
llvm::SmallDenseSet<const ValueDecl *, 4> ImplicitDeclarations;
@@ -3466,6 +3478,19 @@
}
}
}
+ if (SemaRef.getLangOpts().OpenMP > 50) {
+ bool IsModifierPresent = Stack->getDefaultmapModifier(ClauseKind) ==
+ OMPC_DEFAULTMAP_MODIFIER_present;
+ if (IsModifierPresent) {
+ if (llvm::find(ImplicitMapModifier[ClauseKind],
+ OMPC_MAP_MODIFIER_present) ==
+ std::end(ImplicitMapModifier[ClauseKind])) {
+ ImplicitMapModifier[ClauseKind].push_back(
+ OMPC_MAP_MODIFIER_present);
+ ImplicitMapModifierLoc[ClauseKind].push_back(SourceLocation());
+ }
+ }
+ }
if (isOpenMPTargetExecutionDirective(DKind) &&
!Stack->isLoopControlVariable(VD).first) {
@@ -3506,7 +3531,7 @@
Stack->getDefaultmapModifier(ClauseKind);
OpenMPMapClauseKind Kind = getMapClauseKindFromModifier(
M, ClauseKind == OMPC_DEFAULTMAP_aggregate || Res);
- ImplicitMap[Kind].emplace_back(E);
+ ImplicitMap[ClauseKind][Kind].emplace_back(E);
}
return;
}
@@ -3593,9 +3618,11 @@
OpenMPDefaultmapClauseModifier Modifier =
Stack->getDefaultmapModifier(OMPC_DEFAULTMAP_aggregate);
+ OpenMPDefaultmapClauseKind ClauseKind =
+ getVariableCategoryFromDecl(SemaRef.getLangOpts(), FD);
OpenMPMapClauseKind Kind = getMapClauseKindFromModifier(
Modifier, /*IsAggregateOrDeclareTarget*/ true);
- ImplicitMap[Kind].emplace_back(E);
+ ImplicitMap[ClauseKind][Kind].emplace_back(E);
return;
}
@@ -3731,8 +3758,17 @@
ArrayRef<Expr *> getImplicitFirstprivate() const {
return ImplicitFirstprivate;
}
- ArrayRef<Expr *> getImplicitMap(OpenMPDefaultmapClauseKind Kind) const {
- return ImplicitMap[Kind];
+ ArrayRef<Expr *> getImplicitMap(OpenMPDefaultmapClauseKind DK,
+ OpenMPMapClauseKind MK) const {
+ return ImplicitMap[DK][MK];
+ }
+ ArrayRef<OpenMPMapModifierKind>
+ getImplicitMapModifier(OpenMPDefaultmapClauseKind Kind) const {
+ return ImplicitMapModifier[Kind];
+ }
+ ArrayRef<SourceLocation>
+ getImplicitMapModifierLoc(OpenMPDefaultmapClauseKind Kind) const {
+ return ImplicitMapModifierLoc[Kind];
}
const Sema::VarsWithInheritedDSAType &getVarsWithInheritedDSA() const {
return VarsWithInheritedDSA;
@@ -5082,11 +5118,27 @@
SmallVector<Expr *, 4> ImplicitFirstprivates(
DSAChecker.getImplicitFirstprivate().begin(),
DSAChecker.getImplicitFirstprivate().end());
- SmallVector<Expr *, 4> ImplicitMaps[OMPC_MAP_delete];
- for (unsigned I = 0; I < OMPC_MAP_delete; ++I) {
- ArrayRef<Expr *> ImplicitMap =
- DSAChecker.getImplicitMap(static_cast<OpenMPDefaultmapClauseKind>(I));
- ImplicitMaps[I].append(ImplicitMap.begin(), ImplicitMap.end());
+ const unsigned DefaultmapKindNum = OMPC_DEFAULTMAP_pointer + 1;
+ SmallVector<Expr *, 4> ImplicitMaps[DefaultmapKindNum][OMPC_MAP_delete];
+ SmallVector<OpenMPMapModifierKind, NumberOfOMPMapClauseModifiers>
+ ImplicitMapModifiers[DefaultmapKindNum];
+ SmallVector<SourceLocation, NumberOfOMPMapClauseModifiers>
+ ImplicitMapModifiersLoc[DefaultmapKindNum];
+ for (unsigned VC = 0; VC < DefaultmapKindNum; ++VC) {
+ auto Kind = static_cast<OpenMPDefaultmapClauseKind>(VC);
+ for (unsigned I = 0; I < OMPC_MAP_delete; ++I) {
+ ArrayRef<Expr *> ImplicitMap = DSAChecker.getImplicitMap(
+ Kind, static_cast<OpenMPMapClauseKind>(I));
+ ImplicitMaps[VC][I].append(ImplicitMap.begin(), ImplicitMap.end());
+ }
+ ArrayRef<OpenMPMapModifierKind> ImplicitModifier =
+ DSAChecker.getImplicitMapModifier(Kind);
+ ImplicitMapModifiers[VC].append(ImplicitModifier.begin(),
+ ImplicitModifier.end());
+ ArrayRef<SourceLocation> ImplicitModifierLoc =
+ DSAChecker.getImplicitMapModifierLoc(Kind);
+ ImplicitMapModifiersLoc[VC].append(ImplicitModifierLoc.begin(),
+ ImplicitModifierLoc.end());
}
// Mark taskgroup task_reduction descriptors as implicitly firstprivate.
for (OMPClause *C : Clauses) {
@@ -5112,23 +5164,26 @@
ErrorFound = true;
}
}
- int ClauseKindCnt = -1;
- for (ArrayRef<Expr *> ImplicitMap : ImplicitMaps) {
- ++ClauseKindCnt;
- if (ImplicitMap.empty())
- continue;
- CXXScopeSpec MapperIdScopeSpec;
- DeclarationNameInfo MapperId;
- auto Kind = static_cast<OpenMPMapClauseKind>(ClauseKindCnt);
- if (OMPClause *Implicit = ActOnOpenMPMapClause(
- llvm::None, llvm::None, MapperIdScopeSpec, MapperId, Kind,
- /*IsMapTypeImplicit=*/true, SourceLocation(), SourceLocation(),
- ImplicitMap, OMPVarListLocTy())) {
- ClausesWithImplicit.emplace_back(Implicit);
- ErrorFound |=
- cast<OMPMapClause>(Implicit)->varlist_size() != ImplicitMap.size();
- } else {
- ErrorFound = true;
+ for (unsigned I = 0, E = DefaultmapKindNum; I < E; ++I) {
+ int ClauseKindCnt = -1;
+ for (ArrayRef<Expr *> ImplicitMap : ImplicitMaps[I]) {
+ ++ClauseKindCnt;
+ if (ImplicitMap.empty())
+ continue;
+ CXXScopeSpec MapperIdScopeSpec;
+ DeclarationNameInfo MapperId;
+ auto Kind = static_cast<OpenMPMapClauseKind>(ClauseKindCnt);
+ if (OMPClause *Implicit = ActOnOpenMPMapClause(
+ ImplicitMapModifiers[I], ImplicitMapModifiersLoc[I],
+ MapperIdScopeSpec, MapperId, Kind, /*IsMapTypeImplicit=*/true,
+ SourceLocation(), SourceLocation(), ImplicitMap,
+ OMPVarListLocTy())) {
+ ClausesWithImplicit.emplace_back(Implicit);
+ ErrorFound |= cast<OMPMapClause>(Implicit)->varlist_size() !=
+ ImplicitMap.size();
+ } else {
+ ErrorFound = true;
+ }
}
}
}
@@ -18534,20 +18589,38 @@
bool isDefaultmapKind = (Kind != OMPC_DEFAULTMAP_unknown) ||
(LangOpts.OpenMP >= 50 && KindLoc.isInvalid());
if (!isDefaultmapKind || !isDefaultmapModifier) {
- std::string ModifierValue = "'alloc', 'from', 'to', 'tofrom', "
- "'firstprivate', 'none', 'default'";
std::string KindValue = "'scalar', 'aggregate', 'pointer'";
- if (!isDefaultmapKind && isDefaultmapModifier) {
- Diag(KindLoc, diag::err_omp_unexpected_clause_value)
- << KindValue << getOpenMPClauseName(OMPC_defaultmap);
- } else if (isDefaultmapKind && !isDefaultmapModifier) {
- Diag(MLoc, diag::err_omp_unexpected_clause_value)
- << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
+ if (LangOpts.OpenMP == 50) {
+ std::string ModifierValue = "'alloc', 'from', 'to', 'tofrom', "
+ "'firstprivate', 'none', 'default'";
+ if (!isDefaultmapKind && isDefaultmapModifier) {
+ Diag(KindLoc, diag::err_omp_unexpected_clause_value)
+ << KindValue << getOpenMPClauseName(OMPC_defaultmap);
+ } else if (isDefaultmapKind && !isDefaultmapModifier) {
+ Diag(MLoc, diag::err_omp_unexpected_clause_value)
+ << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
+ } else {
+ Diag(MLoc, diag::err_omp_unexpected_clause_value)
+ << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
+ Diag(KindLoc, diag::err_omp_unexpected_clause_value)
+ << KindValue << getOpenMPClauseName(OMPC_defaultmap);
+ }
} else {
- Diag(MLoc, diag::err_omp_unexpected_clause_value)
- << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
- Diag(KindLoc, diag::err_omp_unexpected_clause_value)
- << KindValue << getOpenMPClauseName(OMPC_defaultmap);
+ std::string ModifierValue =
+ "'alloc', 'from', 'to', 'tofrom', "
+ "'firstprivate', 'none', 'default', 'present'";
+ if (!isDefaultmapKind && isDefaultmapModifier) {
+ Diag(KindLoc, diag::err_omp_unexpected_clause_value)
+ << KindValue << getOpenMPClauseName(OMPC_defaultmap);
+ } else if (isDefaultmapKind && !isDefaultmapModifier) {
+ Diag(MLoc, diag::err_omp_unexpected_clause_value)
+ << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
+ } else {
+ Diag(MLoc, diag::err_omp_unexpected_clause_value)
+ << ModifierValue << getOpenMPClauseName(OMPC_defaultmap);
+ Diag(KindLoc, diag::err_omp_unexpected_clause_value)
+ << KindValue << getOpenMPClauseName(OMPC_defaultmap);
+ }
}
return nullptr;
}
Index: clang/include/clang/Basic/OpenMPKinds.def
===================================================================
--- clang/include/clang/Basic/OpenMPKinds.def
+++ clang/include/clang/Basic/OpenMPKinds.def
@@ -89,6 +89,7 @@
OPENMP_DEFAULTMAP_MODIFIER(firstprivate)
OPENMP_DEFAULTMAP_MODIFIER(none)
OPENMP_DEFAULTMAP_MODIFIER(default)
+OPENMP_DEFAULTMAP_MODIFIER(present)
// Static attributes for 'depend' clause.
OPENMP_DEPEND_KIND(in)
_______________________________________________
cfe-commits mailing list
[email protected]
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
