[clang] [OpenMP] Defaultmap: fixes scalar issue, adds all variable category (PR #99315)
via cfe-commits
cfe-commits at lists.llvm.org
Wed Jul 17 05:42:14 PDT 2024
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-clang
Author: None (nicebert)
<details>
<summary>Changes</summary>
Fixes issue with defaultmap where scalar isn't handled correctly for present modifier. Adds all variable cateogry introduced in OpenMP 5.2 and alters existing tests for error messages to check OpenMP 5.2 defaultmap messages.
---
Patch is 117.56 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/99315.diff
13 Files Affected:
- (modified) clang/include/clang/Basic/OpenMPKinds.def (+1)
- (modified) clang/lib/Parse/ParseOpenMP.cpp (+1)
- (modified) clang/lib/Sema/SemaOpenMP.cpp (+14-11)
- (modified) clang/test/OpenMP/target_defaultmap_messages.cpp (+15-12)
- (added) clang/test/OpenMP/target_defaultmap_scalar_codegen.cpp (+71)
- (modified) clang/test/OpenMP/target_parallel_defaultmap_messages.cpp (+13-10)
- (modified) clang/test/OpenMP/target_parallel_for_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_parallel_for_simd_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_simd_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_teams_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_teams_distribute_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_teams_distribute_parallel_for_defaultmap_messages.cpp (+15-12)
- (modified) clang/test/OpenMP/target_teams_distribute_parallel_for_simd_defaultmap_messages.cpp (+15-12)
``````````diff
diff --git a/clang/include/clang/Basic/OpenMPKinds.def b/clang/include/clang/Basic/OpenMPKinds.def
index f46a92d5ecfd4..51084913bf102 100644
--- a/clang/include/clang/Basic/OpenMPKinds.def
+++ b/clang/include/clang/Basic/OpenMPKinds.def
@@ -107,6 +107,7 @@ OPENMP_DEVICE_MODIFIER(device_num)
OPENMP_DEFAULTMAP_KIND(scalar)
OPENMP_DEFAULTMAP_KIND(aggregate)
OPENMP_DEFAULTMAP_KIND(pointer)
+OPENMP_DEFAULTMAP_KIND(all)
// Modifiers for 'defaultmap' clause.
OPENMP_DEFAULTMAP_MODIFIER(alloc)
diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp
index 326cd22ff9005..35f37139b0064 100644
--- a/clang/lib/Parse/ParseOpenMP.cpp
+++ b/clang/lib/Parse/ParseOpenMP.cpp
@@ -3835,6 +3835,7 @@ OMPClause *Parser::ParseOpenMPSingleExprWithArgClause(OpenMPDirectiveKind DKind,
// Get a defaultmap modifier
unsigned Modifier = getOpenMPSimpleClauseType(
Kind, Tok.isAnnotation() ? "" : PP.getSpelling(Tok), getLangOpts());
+
// Set defaultmap modifier to unknown if it is either scalar, aggregate, or
// pointer
if (Modifier < OMPC_DEFAULTMAP_MODIFIER_unknown)
diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp
index bc03280aa8aaf..9d0746ea9be48 100644
--- a/clang/lib/Sema/SemaOpenMP.cpp
+++ b/clang/lib/Sema/SemaOpenMP.cpp
@@ -822,7 +822,8 @@ class DSAStackTy {
return (M == OMPC_DEFAULTMAP_MODIFIER_alloc) ||
(M == OMPC_DEFAULTMAP_MODIFIER_to) ||
(M == OMPC_DEFAULTMAP_MODIFIER_from) ||
- (M == OMPC_DEFAULTMAP_MODIFIER_tofrom);
+ (M == OMPC_DEFAULTMAP_MODIFIER_tofrom) ||
+ (M == OMPC_DEFAULTMAP_MODIFIER_present);
}
return true;
}
@@ -3099,11 +3100,11 @@ ExprResult SemaOpenMP::ActOnOpenMPIdExpression(Scope *CurScope,
if (TypoCorrection Corrected =
SemaRef.CorrectTypo(Id, Sema::LookupOrdinaryName, CurScope, nullptr,
CCC, Sema::CTK_ErrorRecovery)) {
- SemaRef.diagnoseTypo(Corrected,
- PDiag(Lookup.empty()
- ? diag::err_undeclared_var_use_suggest
- : diag::err_omp_expected_var_arg_suggest)
- << Id.getName());
+ SemaRef.diagnoseTypo(
+ Corrected,
+ SemaRef.PDiag(Lookup.empty() ? diag::err_undeclared_var_use_suggest
+ : diag::err_omp_expected_var_arg_suggest)
+ << Id.getName());
VD = Corrected.getCorrectionDeclAs<VarDecl>();
} else {
Diag(Id.getLoc(), Lookup.empty() ? diag::err_undeclared_var_use
@@ -7591,9 +7592,9 @@ SemaOpenMP::checkOpenMPDeclareVariantFunction(SemaOpenMP::DeclGroupPtrTy DG,
PartialDiagnostic::NullDiagnostic()),
PartialDiagnosticAt(
VariantRef->getExprLoc(),
- PDiag(diag::err_omp_declare_variant_doesnt_support)),
+ SemaRef.PDiag(diag::err_omp_declare_variant_doesnt_support)),
PartialDiagnosticAt(VariantRef->getExprLoc(),
- PDiag(diag::err_omp_declare_variant_diff)
+ SemaRef.PDiag(diag::err_omp_declare_variant_diff)
<< FD->getLocation()),
/*TemplatesSupported=*/true, /*ConstexprSupported=*/false,
/*CLinkageMayDiffer=*/true))
@@ -21665,7 +21666,9 @@ OMPClause *SemaOpenMP::ActOnOpenMPDefaultmapClause(
bool isDefaultmapKind = (Kind != OMPC_DEFAULTMAP_unknown) ||
(getLangOpts().OpenMP >= 50 && KindLoc.isInvalid());
if (!isDefaultmapKind || !isDefaultmapModifier) {
- StringRef KindValue = "'scalar', 'aggregate', 'pointer'";
+ StringRef KindValue = getLangOpts().OpenMP < 52
+ ? "'scalar', 'aggregate', 'pointer'"
+ : "'scalar', 'aggregate', 'pointer', 'all'";
if (getLangOpts().OpenMP == 50) {
StringRef ModifierValue = "'alloc', 'from', 'to', 'tofrom', "
"'firstprivate', 'none', 'default'";
@@ -21709,7 +21712,7 @@ OMPClause *SemaOpenMP::ActOnOpenMPDefaultmapClause(
return nullptr;
}
}
- if (Kind == OMPC_DEFAULTMAP_unknown) {
+ if (Kind == OMPC_DEFAULTMAP_unknown || Kind == OMPC_DEFAULTMAP_all) {
// Variable category is not specified - mark all categories.
DSAStack->setDefaultDMAAttr(M, OMPC_DEFAULTMAP_aggregate, StartLoc);
DSAStack->setDefaultDMAAttr(M, OMPC_DEFAULTMAP_scalar, StartLoc);
@@ -21782,7 +21785,7 @@ NamedDecl *SemaOpenMP::lookupOpenMPDeclareTargetName(
SemaRef.CorrectTypo(Id, Sema::LookupOrdinaryName, CurScope, nullptr,
CCC, Sema::CTK_ErrorRecovery)) {
SemaRef.diagnoseTypo(Corrected,
- PDiag(diag::err_undeclared_var_use_suggest)
+ SemaRef.PDiag(diag::err_undeclared_var_use_suggest)
<< Id.getName());
checkDeclIsAllowedInOpenMPTarget(nullptr, Corrected.getCorrectionDecl());
return nullptr;
diff --git a/clang/test/OpenMP/target_defaultmap_messages.cpp b/clang/test/OpenMP/target_defaultmap_messages.cpp
index 8052e34d5c933..88ae3b7962d55 100644
--- a/clang/test/OpenMP/target_defaultmap_messages.cpp
+++ b/clang/test/OpenMP/target_defaultmap_messages.cpp
@@ -1,5 +1,8 @@
-// RUN: %clang_cc1 -verify -Wno-vla -fopenmp %s -verify=expected,omp51 -Wuninitialized -DOMP51
-// RUN: %clang_cc1 -verify -Wno-vla -fopenmp-simd %s -verify=expected,omp51 -Wuninitialized -DOMP51
+// RUN: %clang_cc1 -verify -Wno-vla -fopenmp %s -fopenmp-version=52 -verify=expected,omp5x,omp52 -Wuninitialized
+// RUN: %clang_cc1 -verify -Wno-vla -fopenmp-simd %s -fopenmp-version=52 -verify=expected,omp5x,omp52 -Wuninitialized
+
+// RUN: %clang_cc1 -verify -Wno-vla -fopenmp %s -fopenmp-version=51 -verify=expected,omp5x,omp51 -Wuninitialized -DOMP51
+// RUN: %clang_cc1 -verify -Wno-vla -fopenmp-simd %s -fopenmp-version=51 -verify=expected,omp5x,omp51 -Wuninitialized -DOMP51
// RUN: %clang_cc1 -verify -Wno-vla -fopenmp -fopenmp-version=50 %s -verify=expected,omp5 -Wuninitialized -DOMP5
// RUN: %clang_cc1 -verify -Wno-vla -fopenmp-simd -fopenmp-version=50 %s -verify=expected,omp5 -Wuninitialized -DOMP5
@@ -30,23 +33,23 @@ template <class T, typename S, int N, int ST>
T tmain(T argc, S **argv) {
#pragma omp target defaultmap // expected-error {{expected '(' after 'defaultmap'}}
foo();
-#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'}}
+#pragma omp target defaultmap( // omp5x-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() // 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'}}
+#pragma omp target defaultmap() // omp5x-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 '('}} 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'}}
+ #pragma omp target defaultmap (tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp52-error {{expected 'scalar', 'aggregate', 'pointer', 'all' in OpenMP clause 'defaultmap'}} 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: // 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'}}
+ #pragma omp target defaultmap (scalar: // omp5x-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp52-error {{expected 'scalar', 'aggregate', 'pointer', 'all' 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}} omp51-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}} omp5x-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
foo();
@@ -93,23 +96,23 @@ T tmain(T argc, S **argv) {
int main(int argc, char **argv) {
#pragma omp target defaultmap // expected-error {{expected '(' after 'defaultmap'}}
foo();
-#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'}}
+#pragma omp target defaultmap( // omp5x-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() // 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'}}
+#pragma omp target defaultmap() // omp5x-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 '('}} 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'}}
+#pragma omp target defaultmap(tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp52-error {{expected 'scalar', 'aggregate', 'pointer', 'all' in OpenMP clause 'defaultmap'}} 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: // 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'}}
+#pragma omp target defaultmap(scalar: // omp5x-error {{expected 'alloc', 'from', 'to', 'tofrom', 'firstprivate', 'none', 'default', 'present' in OpenMP clause 'defaultmap'}} omp52-error {{expected 'scalar', 'aggregate', 'pointer', 'all' 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}} omp51-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}} omp5x-error {{at most one defaultmap clause for each variable-category can appear on the directive}}
foo();
#ifdef OMP5
diff --git a/clang/test/OpenMP/target_defaultmap_scalar_codegen.cpp b/clang/test/OpenMP/target_defaultmap_scalar_codegen.cpp
new file mode 100644
index 0000000000000..75274683d4c70
--- /dev/null
+++ b/clang/test/OpenMP/target_defaultmap_scalar_codegen.cpp
@@ -0,0 +1,71 @@
+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -emit-llvm -o - -fopenmp \
+// RUN: -triple x86_64-unknown-linux-gnu %s | FileCheck %s
+
+// expected-no-diagnostics
+#ifndef HEADER
+#define HEADER
+
+int main() {
+ int scalar_var = 1; // scalar
+ int errors = 0;
+
+ #pragma omp target enter data map(to: scalar_var)
+
+ #pragma omp target map(tofrom: errors) defaultmap(present)
+ {
+ if(scalar_var != 1) errors++;
+
+ scalar_var = 7;
+ }
+
+ if(scalar_var != 1) errors++;
+
+ #pragma omp target exit data map(delete: scalar_var)
+}
+
+#endif
+// CHECK-LABEL: define {{[^@]+}}@main
+// CHECK-SAME: () #[[ATTR0:[0-9]+]] {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[RETVAL:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[SCALAR_VAR:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[ERRORS:%.*]] = alloca i32, align 4
+// CHECK-NEXT: store i32 0, ptr [[RETVAL]], align 4
+// CHECK-NEXT: store i32 1, ptr [[SCALAR_VAR]], align 4
+// CHECK-NEXT: store i32 0, ptr [[ERRORS]], align 4
+// CHECK-NEXT: call void @__omp_offloading_fd00_3464d5f_main_l37(ptr [[SCALAR_VAR]], ptr [[ERRORS]]) #[[ATTR2:[0-9]+]]
+// CHECK-NEXT: [[TMP0:%.*]] = load i32, ptr [[SCALAR_VAR]], align 4
+// CHECK-NEXT: [[CMP:%.*]] = icmp ne i32 [[TMP0]], 1
+// CHECK-NEXT: br i1 [[CMP]], label [[IF_THEN:%.*]], label [[IF_END:%.*]]
+// CHECK: if.then:
+// CHECK-NEXT: [[TMP1:%.*]] = load i32, ptr [[ERRORS]], align 4
+// CHECK-NEXT: [[INC:%.*]] = add nsw i32 [[TMP1]], 1
+// CHECK-NEXT: store i32 [[INC]], ptr [[ERRORS]], align 4
+// CHECK-NEXT: br label [[IF_END]]
+// CHECK: if.end:
+// CHECK-NEXT: [[TMP2:%.*]] = load i32, ptr [[RETVAL]], align 4
+// CHECK-NEXT: ret i32 [[TMP2]]
+//
+//
+// CHECK-LABEL: define {{[^@]+}}@__omp_offloading_fd00_3464d5f_main_l37
+// CHECK-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[SCALAR_VAR:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[ERRORS:%.*]]) #[[ATTR1:[0-9]+]] {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[SCALAR_VAR_ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: [[ERRORS_ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: store ptr [[SCALAR_VAR]], ptr [[SCALAR_VAR_ADDR]], align 8
+// CHECK-NEXT: store ptr [[ERRORS]], ptr [[ERRORS_ADDR]], align 8
+// CHECK-NEXT: [[TMP0:%.*]] = load ptr, ptr [[SCALAR_VAR_ADDR]], align 8
+// CHECK-NEXT: [[TMP1:%.*]] = load ptr, ptr [[ERRORS_ADDR]], align 8
+// CHECK-NEXT: [[TMP2:%.*]] = load i32, ptr [[TMP0]], align 4
+// CHECK-NEXT: [[CMP:%.*]] = icmp ne i32 [[TMP2]], 1
+// CHECK-NEXT: br i1 [[CMP]...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/99315
More information about the cfe-commits
mailing list