jdenny created this revision.
jdenny added a reviewer: ABataev.
Herald added subscribers: jdoerfert, guansong.
Herald added a project: clang.
For example, without this fix, when the host is x86_64, long double is
sometimes rejected when offloading to x86_64:
$ cat test.c
int main() {
long double in = 0;
#pragma omp target
in *= 2;
return 0;
}
$ clang -fopenmp test.c; echo $?
0
$ clang -fopenmp -fopenmp-targets=x86_64 test.c
test.c:4:5: error: 'long double' is not supported on this target
in *= 2;
^~
1 error generated.
Repository:
rG LLVM Github Monorepo
https://reviews.llvm.org/D64289
Files:
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
Index: clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
===================================================================
--- clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
+++ clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
@@ -1,8 +1,16 @@
// Test target codegen - host bc file has to be created first.
-// RUN: %clang_cc1 -fopenmp -x c++ -triple x86_64-unknown-linux
-fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple x86_64-unknown-linux
-fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown
-aux-triple x86_64-unknown-linux -fopenmp-targets=nvptx64-nvidia-cuda %s
-fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
-// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu
-fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple
powerpc64le-unknown-linux-gnu -fopenmp-targets=nvptx64-nvidia-cuda
-emit-llvm-bc %s -o %t-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown
-aux-triple powerpc64le-unknown-linux-gnu -fopenmp-targets=nvptx64-nvidia-cuda
%s -fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+//
+// Make sure this code does compile when the host and target are the same.
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple x86_64-unknown-linux
-fopenmp-targets=x86_64-unknown-linux -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple
powerpc64le-unknown-linux-gnu -aux-triple x86_64-unknown-linux
-fopenmp-targets=x86_64-unknown-linux %s -fopenmp-is-device
-fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple
powerpc64le-unknown-linux-gnu -fopenmp-targets=powerpc64le-unknown-linux-gnu
-emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple
powerpc64le-unknown-linux-gnu -aux-triple powerpc64le-unknown-linux-gnu
-fopenmp-targets=powerpc64le-unknown-linux-gnu %s -fopenmp-is-device
-fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+
+// quiet-no-diagnostics
struct T {
char a;
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -1590,7 +1590,8 @@
if ((Ty->isFloat16Type() && !Context.getTargetInfo().hasFloat16Type()) ||
((Ty->isFloat128Type() ||
(Ty->isRealFloatingType() && Context.getTypeSize(Ty) == 128)) &&
- !Context.getTargetInfo().hasFloat128Type()) ||
+ !Context.getTargetInfo().hasFloat128Type() &&
+ Context.getTargetInfo().getLongDoubleWidth() != 128) ||
(Ty->isIntegerType() && Context.getTypeSize(Ty) == 128 &&
!Context.getTargetInfo().hasInt128Type()))
targetDiag(E->getExprLoc(), diag::err_type_unsupported)
Index: clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
===================================================================
--- clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
+++ clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
@@ -1,8 +1,16 @@
// Test target codegen - host bc file has to be created first.
-// RUN: %clang_cc1 -fopenmp -x c++ -triple x86_64-unknown-linux -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple x86_64-unknown-linux -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple x86_64-unknown-linux -fopenmp-targets=nvptx64-nvidia-cuda %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
-// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-linux-gnu -fopenmp-targets=nvptx64-nvidia-cuda %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+//
+// Make sure this code does compile when the host and target are the same.
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple x86_64-unknown-linux -fopenmp-targets=x86_64-unknown-linux -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu -aux-triple x86_64-unknown-linux -fopenmp-targets=x86_64-unknown-linux %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu -fopenmp-targets=powerpc64le-unknown-linux-gnu -emit-llvm-bc %s -o %t-host.bc
+// RUN: %clang_cc1 -verify=quiet -fopenmp -x c++ -triple powerpc64le-unknown-linux-gnu -aux-triple powerpc64le-unknown-linux-gnu -fopenmp-targets=powerpc64le-unknown-linux-gnu %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-host.bc -fsyntax-only
+
+// quiet-no-diagnostics
struct T {
char a;
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -1590,7 +1590,8 @@
if ((Ty->isFloat16Type() && !Context.getTargetInfo().hasFloat16Type()) ||
((Ty->isFloat128Type() ||
(Ty->isRealFloatingType() && Context.getTypeSize(Ty) == 128)) &&
- !Context.getTargetInfo().hasFloat128Type()) ||
+ !Context.getTargetInfo().hasFloat128Type() &&
+ Context.getTargetInfo().getLongDoubleWidth() != 128) ||
(Ty->isIntegerType() && Context.getTypeSize(Ty) == 128 &&
!Context.getTargetInfo().hasInt128Type()))
targetDiag(E->getExprLoc(), diag::err_type_unsupported)
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
