jhuber6 updated this revision to Diff 420604.
jhuber6 added a comment.
Updating to stop erroring if a function with the same mangled error exists, but
is never emitted.
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D122760/new/
https://reviews.llvm.org/D122760
Files:
clang/include/clang/Basic/Attr.td
clang/lib/CodeGen/CodeGenModule.cpp
clang/lib/CodeGen/CodeGenModule.h
clang/lib/Parse/ParseOpenMP.cpp
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/declare_variant_no_mangling.cpp
clang/test/OpenMP/nvptx_declare_variant_name_mangling.cpp
llvm/include/llvm/Frontend/OpenMP/OMPKinds.def
Index: llvm/include/llvm/Frontend/OpenMP/OMPKinds.def
===================================================================
--- llvm/include/llvm/Frontend/OpenMP/OMPKinds.def
+++ llvm/include/llvm/Frontend/OpenMP/OMPKinds.def
@@ -1157,6 +1157,7 @@
__OMP_TRAIT_PROPERTY(implementation, extension, match_none)
__OMP_TRAIT_PROPERTY(implementation, extension, disable_implicit_base)
__OMP_TRAIT_PROPERTY(implementation, extension, allow_templates)
+__OMP_TRAIT_PROPERTY(implementation, extension, keep_original_name)
__OMP_TRAIT_SET(user)
Index: clang/test/OpenMP/nvptx_declare_variant_name_mangling.cpp
===================================================================
--- clang/test/OpenMP/nvptx_declare_variant_name_mangling.cpp
+++ clang/test/OpenMP/nvptx_declare_variant_name_mangling.cpp
@@ -1,13 +1,15 @@
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
-// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --implicit-check-not='call i32 {@_Z3bazv|@_Z3barv}'
+// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --implicit-check-not='call i32 {@_Z3bazv|@_Z3barv|@_ZL3foov}'
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
-// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - | FileCheck %s --implicit-check-not='call i32 {@_Z3bazv|@_Z3barv}'
+// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - | FileCheck %s --implicit-check-not='call i32 {@_Z3bazv|@_Z3barv|@_ZL3foov}'
// expected-no-diagnostics
// CHECK-DAG: @_Z3barv
// CHECK-DAG: @_Z3bazv
+// CHECK-DAG: call noundef i32 @_ZL3foov()
// CHECK-DAG: define{{.*}} @"_Z53bar$ompvariant$S2$s7$Pnvptx$Pnvptx64$S3$s9$Pmatch_anyv"
// CHECK-DAG: define{{.*}} @"_Z53baz$ompvariant$S2$s7$Pnvptx$Pnvptx64$S3$s9$Pmatch_anyv"
+// CHECK-DAG: define{{.*}} @_ZL3foov
// CHECK-DAG: call noundef i32 @"_Z53bar$ompvariant$S2$s7$Pnvptx$Pnvptx64$S3$s9$Pmatch_anyv"()
// CHECK-DAG: call noundef i32 @"_Z53baz$ompvariant$S2$s7$Pnvptx$Pnvptx64$S3$s9$Pmatch_anyv"()
@@ -20,6 +22,8 @@
int baz() { return 5; }
+static int foo() { return 3; }
+
#pragma omp begin declare variant match(device = {arch(nvptx, nvptx64)}, implementation = {extension(match_any)})
int bar() { return 2; }
@@ -28,13 +32,19 @@
#pragma omp end declare variant
+#pragma omp begin declare variant match(device = {arch(nvptx, nvptx64)}, implementation = {extension(match_any, keep_original_name)})
+
+static int foo() { return 4; }
+
+#pragma omp end declare variant
+
#pragma omp end declare target
int main() {
int res;
#pragma omp target map(from \
: res)
- res = bar() + baz();
+ res = bar() + baz() + foo();
return res;
}
Index: clang/test/OpenMP/declare_variant_no_mangling.cpp
===================================================================
--- /dev/null
+++ clang/test/OpenMP/declare_variant_no_mangling.cpp
@@ -0,0 +1,53 @@
+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+static int foo() { return 0; }
+extern inline int bar() { return 0; }
+int baz() { return 0; }
+
+#pragma omp begin declare variant match( \
+ device = {arch(ppc64le, ppc64)}, \
+ implementation = {extension(match_any, keep_original_name)})
+
+static int foo() { return 1; }
+extern inline int bar() { return 1; }
+
+#pragma omp end declare variant
+
+int main() {
+ return foo() + bar();
+}
+
+#endif
+// CHECK-LABEL: define {{[^@]+}}@_Z3bazv
+// CHECK-SAME: () #[[ATTR0:[0-9]+]] {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: ret i32 0
+//
+//
+// CHECK-LABEL: define {{[^@]+}}@main
+// CHECK-SAME: () #[[ATTR1:[0-9]+]] {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[RETVAL:%.*]] = alloca i32, align 4
+// CHECK-NEXT: store i32 0, i32* [[RETVAL]], align 4
+// CHECK-NEXT: [[CALL:%.*]] = call noundef signext i32 @_ZL3foov()
+// CHECK-NEXT: [[CALL1:%.*]] = call noundef signext i32 @_Z3barv()
+// CHECK-NEXT: [[ADD:%.*]] = add nsw i32 [[CALL]], [[CALL1]]
+// CHECK-NEXT: ret i32 [[ADD]]
+//
+//
+// CHECK-LABEL: define {{[^@]+}}@_ZL3foov
+// CHECK-SAME: () #[[ATTR0]] {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: ret i32 1
+//
+//
+// CHECK-LABEL: define {{[^@]+}}@_Z3barv
+// CHECK-SAME: () #[[ATTR0]] comdat {
+// CHECK-NEXT: entry:
+// CHECK-NEXT: ret i32 1
+//
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -6910,6 +6910,11 @@
/*AppendArgs=*/nullptr, /*AppendArgsSize=*/0);
for (FunctionDecl *BaseFD : Bases)
BaseFD->addAttr(OMPDeclareVariantA);
+ if (DVScope.TI->isExtensionActive(
+ llvm::omp::TraitProperty::
+ implementation_extension_keep_original_name))
+ D->addAttr(
+ OMPDeclareVariantNoMangleAttr::CreateImplicit(Context, Bases.front()));
}
ExprResult Sema::ActOnOpenMPCall(ExprResult Call, Scope *Scope,
Index: clang/lib/Parse/ParseOpenMP.cpp
===================================================================
--- clang/lib/Parse/ParseOpenMP.cpp
+++ clang/lib/Parse/ParseOpenMP.cpp
@@ -957,6 +957,10 @@
TraitProperty::implementation_extension_allow_templates)
return true;
+ if (TIProperty.Kind ==
+ TraitProperty::implementation_extension_keep_original_name)
+ return true;
+
auto IsMatchExtension = [](OMPTraitProperty &TP) {
return (TP.Kind ==
llvm::omp::TraitProperty::implementation_extension_match_all ||
Index: clang/lib/CodeGen/CodeGenModule.h
===================================================================
--- clang/lib/CodeGen/CodeGenModule.h
+++ clang/lib/CodeGen/CodeGenModule.h
@@ -370,6 +370,10 @@
/// for the same decl.
llvm::DenseSet<GlobalDecl> DiagnosedConflictingDefinitions;
+ /// Set of OpenMP declare variant globals actually emitted. Required to not
+ /// issue a mangling conflict on a variant function that is never emitted.
+ llvm::DenseSet<GlobalDecl> VariantGlobalsEmitted;
+
/// A queue of (optional) vtables to consider emitting.
std::vector<const CXXRecordDecl*> DeferredVTables;
Index: clang/lib/CodeGen/CodeGenModule.cpp
===================================================================
--- clang/lib/CodeGen/CodeGenModule.cpp
+++ clang/lib/CodeGen/CodeGenModule.cpp
@@ -1455,6 +1455,8 @@
// Keep the first result in the case of a mangling collision.
const auto *ND = cast<NamedDecl>(GD.getDecl());
+ if (auto *A = ND->getAttr<OMPDeclareVariantNoMangleAttr>())
+ ND = A->getFunction();
std::string MangledName = getMangledNameImpl(*this, GD, ND);
// Ensure either we have different ABIs between host and device compilations,
@@ -3053,6 +3055,11 @@
EmitOMPDeclareMapper(DMD);
return;
}
+
+ // If this variant should keep its original name, mark the original
+ // declaration so we can make sure we don't conflict with it.
+ if (auto *A = Global->getAttr<OMPDeclareVariantNoMangleAttr>())
+ VariantGlobalsEmitted.insert(A->getFunction());
}
// Ignore declarations, they will be emitted on their first use.
@@ -3794,7 +3801,9 @@
if (lookupRepresentativeDecl(MangledName, OtherGD) &&
(GD.getCanonicalDecl().getDecl() !=
OtherGD.getCanonicalDecl().getDecl()) &&
+ VariantGlobalsEmitted.contains(OtherGD) &&
DiagnosedConflictingDefinitions.insert(GD).second) {
+
getDiags().Report(D->getLocation(), diag::err_duplicate_mangled_name)
<< MangledName;
getDiags().Report(OtherGD.getDecl()->getLocation(),
@@ -4112,7 +4121,7 @@
if (D && lookupRepresentativeDecl(MangledName, OtherGD) &&
(D->getCanonicalDecl() != OtherGD.getCanonicalDecl().getDecl()) &&
(OtherD = dyn_cast<VarDecl>(OtherGD.getDecl())) &&
- OtherD->hasInit() &&
+ OtherD->hasInit() && VariantGlobalsEmitted.contains(OtherGD) &&
DiagnosedConflictingDefinitions.insert(D).second) {
getDiags().Report(D->getLocation(), diag::err_duplicate_mangled_name)
<< MangledName;
Index: clang/include/clang/Basic/Attr.td
===================================================================
--- clang/include/clang/Basic/Attr.td
+++ clang/include/clang/Basic/Attr.td
@@ -3786,6 +3786,14 @@
}];
}
+def OMPDeclareVariantNoMangle : InheritableAttr {
+ // This attribute has no spellings as it is only ever created implicitly.
+ let Spellings = [];
+ let Args = [DeclArgument<Function, "Function">];
+ let Subjects = SubjectList<[Function]>;
+ let Documentation = [Undocumented];
+}
+
def Assumption : InheritableAttr {
let Spellings = [Clang<"assume">];
let Subjects = SubjectList<[Function, ObjCMethod]>;
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
