[clang] 06621ec - [Clang] Convert some tests to opaque pointers (NFC)
Nikita Popov via cfe-commits
cfe-commits at lists.llvm.org
Fri Feb 17 06:09:18 PST 2023
Author: Nikita Popov
Date: 2023-02-17T15:08:50+01:00
New Revision: 06621ecdaf416d87968dcaf57a634c83cf0dc1b8
URL: https://github.com/llvm/llvm-project/commit/06621ecdaf416d87968dcaf57a634c83cf0dc1b8
DIFF: https://github.com/llvm/llvm-project/commit/06621ecdaf416d87968dcaf57a634c83cf0dc1b8.diff
LOG: [Clang] Convert some tests to opaque pointers (NFC)
Added:
Modified:
clang/test/CodeGen/math-builtins.c
clang/test/CodeGen/math-libcalls.c
clang/test/CodeGen/mult-alt-generic.c
clang/test/CodeGenCUDA/device-stub.cu
clang/test/CodeGenCUDA/static-device-var-rdc.cu
clang/test/CodeGenCXX/auto-var-init-stop-after.cpp
Removed:
################################################################################
diff --git a/clang/test/CodeGen/math-builtins.c b/clang/test/CodeGen/math-builtins.c
index 559421a4882c5..04738cce01073 100644
--- a/clang/test/CodeGen/math-builtins.c
+++ b/clang/test/CodeGen/math-builtins.c
@@ -1,7 +1,7 @@
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -w -S -o - -emit-llvm %s | FileCheck %s -check-prefix=NO__ERRNO
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s -check-prefix=HAS_ERRNO
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown-gnu -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_GNU
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-windows-msvc -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_WIN
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -w -S -o - -emit-llvm %s | FileCheck %s -check-prefix=NO__ERRNO
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s -check-prefix=HAS_ERRNO
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown-gnu -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_GNU
+// RUN: %clang_cc1 -triple x86_64-unknown-windows-msvc -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_WIN
// Test attributes and codegen of math builtins.
@@ -52,14 +52,14 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) {
__builtin_frexp(f,i); __builtin_frexpf(f,i); __builtin_frexpl(f,i); __builtin_frexpf128(f,i);
-// NO__ERRNO: declare double @frexp(double noundef, i32* noundef) [[NOT_READNONE:#[0-9]+]]
-// NO__ERRNO: declare float @frexpf(float noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare fp128 @frexpf128(fp128 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare double @frexp(double noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare float @frexpf(float noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare fp128 @frexpf128(fp128 noundef, i32* noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare double @frexp(double noundef, ptr noundef) [[NOT_READNONE:#[0-9]+]]
+// NO__ERRNO: declare float @frexpf(float noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare fp128 @frexpf128(fp128 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare double @frexp(double noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare float @frexpf(float noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare fp128 @frexpf128(fp128 noundef, ptr noundef) [[NOT_READNONE]]
__builtin_huge_val(); __builtin_huge_valf(); __builtin_huge_vall(); __builtin_huge_valf128();
@@ -88,36 +88,36 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) {
__builtin_modf(f,d); __builtin_modff(f,fp); __builtin_modfl(f,l); __builtin_modff128(f,l);
-// NO__ERRNO: declare double @modf(double noundef, double* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare float @modff(float noundef, float* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, x86_fp80* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare fp128 @modff128(fp128 noundef, fp128* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare double @modf(double noundef, double* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare float @modff(float noundef, float* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, x86_fp80* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare fp128 @modff128(fp128 noundef, fp128* noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare double @modf(double noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare float @modff(float noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare fp128 @modff128(fp128 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare double @modf(double noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare float @modff(float noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare fp128 @modff128(fp128 noundef, ptr noundef) [[NOT_READNONE]]
__builtin_nan(c); __builtin_nanf(c); __builtin_nanl(c); __builtin_nanf128(c);
-// NO__ERRNO: declare double @nan(i8* noundef) [[PURE:#[0-9]+]]
-// NO__ERRNO: declare float @nanf(i8* noundef) [[PURE]]
-// NO__ERRNO: declare x86_fp80 @nanl(i8* noundef) [[PURE]]
-// NO__ERRNO: declare fp128 @nanf128(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare double @nan(i8* noundef) [[PURE:#[0-9]+]]
-// HAS_ERRNO: declare float @nanf(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare x86_fp80 @nanl(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare fp128 @nanf128(i8* noundef) [[PURE]]
+// NO__ERRNO: declare double @nan(ptr noundef) [[PURE:#[0-9]+]]
+// NO__ERRNO: declare float @nanf(ptr noundef) [[PURE]]
+// NO__ERRNO: declare x86_fp80 @nanl(ptr noundef) [[PURE]]
+// NO__ERRNO: declare fp128 @nanf128(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare double @nan(ptr noundef) [[PURE:#[0-9]+]]
+// HAS_ERRNO: declare float @nanf(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare x86_fp80 @nanl(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare fp128 @nanf128(ptr noundef) [[PURE]]
__builtin_nans(c); __builtin_nansf(c); __builtin_nansl(c); __builtin_nansf128(c);
-// NO__ERRNO: declare double @nans(i8* noundef) [[PURE]]
-// NO__ERRNO: declare float @nansf(i8* noundef) [[PURE]]
-// NO__ERRNO: declare x86_fp80 @nansl(i8* noundef) [[PURE]]
-// NO__ERRNO: declare fp128 @nansf128(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare double @nans(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare float @nansf(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare x86_fp80 @nansl(i8* noundef) [[PURE]]
-// HAS_ERRNO: declare fp128 @nansf128(i8* noundef) [[PURE]]
+// NO__ERRNO: declare double @nans(ptr noundef) [[PURE]]
+// NO__ERRNO: declare float @nansf(ptr noundef) [[PURE]]
+// NO__ERRNO: declare x86_fp80 @nansl(ptr noundef) [[PURE]]
+// NO__ERRNO: declare fp128 @nansf128(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare double @nans(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare float @nansf(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare x86_fp80 @nansl(ptr noundef) [[PURE]]
+// HAS_ERRNO: declare fp128 @nansf128(ptr noundef) [[PURE]]
__builtin_pow(f,f); __builtin_powf(f,f); __builtin_powl(f,f); __builtin_powf128(f,f);
@@ -549,14 +549,14 @@ __builtin_remainder(f,f); __builtin_remainderf(f,f); __builtin_remainderl(f,f);
__builtin_remquo(f,f,i); __builtin_remquof(f,f,i); __builtin_remquol(f,f,i); __builtin_remquof128(f,f,i);
-// NO__ERRNO: declare double @remquo(double noundef, double noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare float @remquof(float noundef, float noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare fp128 @remquof128(fp128 noundef, fp128 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare double @remquo(double noundef, double noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare float @remquof(float noundef, float noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare fp128 @remquof128(fp128 noundef, fp128 noundef, i32* noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare double @remquo(double noundef, double noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare float @remquof(float noundef, float noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare fp128 @remquof128(fp128 noundef, fp128 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare double @remquo(double noundef, double noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare float @remquof(float noundef, float noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare fp128 @remquof128(fp128 noundef, fp128 noundef, ptr noundef) [[NOT_READNONE]]
__builtin_rint(f); __builtin_rintf(f); __builtin_rintl(f); __builtin_rintf128(f);
diff --git a/clang/test/CodeGen/math-libcalls.c b/clang/test/CodeGen/math-libcalls.c
index 0e61f92f8c751..fa8f49d8a2c9f 100644
--- a/clang/test/CodeGen/math-libcalls.c
+++ b/clang/test/CodeGen/math-libcalls.c
@@ -1,8 +1,8 @@
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm %s | FileCheck %s --check-prefix=NO__ERRNO
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm -ffp-exception-behavior=maytrap %s | FileCheck %s --check-prefix=HAS_MAYTRAP
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown-gnu -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_GNU
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-windows-msvc -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_WIN
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm %s | FileCheck %s --check-prefix=NO__ERRNO
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -Wno-implicit-function-declaration -w -S -o - -emit-llvm -ffp-exception-behavior=maytrap %s | FileCheck %s --check-prefix=HAS_MAYTRAP
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown-gnu -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_GNU
+// RUN: %clang_cc1 -triple x86_64-unknown-windows-msvc -Wno-implicit-function-declaration -w -S -o - -emit-llvm -fmath-errno %s | FileCheck %s --check-prefix=HAS_ERRNO_WIN
// Test attributes and builtin codegen of math library calls.
@@ -57,15 +57,15 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) {
frexp(f,i); frexpf(f,i); frexpl(f,i);
- // NO__ERRNO: declare double @frexp(double noundef, i32* noundef) [[NOT_READNONE:#[0-9]+]]
- // NO__ERRNO: declare float @frexpf(float noundef, i32* noundef) [[NOT_READNONE]]
- // NO__ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare double @frexp(double noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare float @frexpf(float noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare double @frexp(double noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare float @frexpf(float noundef, i32* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare x86_fp80 @frexpl(x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
+ // NO__ERRNO: declare double @frexp(double noundef, ptr noundef) [[NOT_READNONE:#[0-9]+]]
+ // NO__ERRNO: declare float @frexpf(float noundef, ptr noundef) [[NOT_READNONE]]
+ // NO__ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare double @frexp(double noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare float @frexpf(float noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare x86_fp80 @frexpl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare double @frexp(double noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare float @frexpf(float noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare x86_fp80 @frexpl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
ldexp(f,f); ldexpf(f,f); ldexpl(f,f);
@@ -81,27 +81,27 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) {
modf(f,d); modff(f,fp); modfl(f,l);
- // NO__ERRNO: declare double @modf(double noundef, double* noundef) [[NOT_READNONE]]
- // NO__ERRNO: declare float @modff(float noundef, float* noundef) [[NOT_READNONE]]
- // NO__ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, x86_fp80* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare double @modf(double noundef, double* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare float @modff(float noundef, float* noundef) [[NOT_READNONE]]
- // HAS_ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, x86_fp80* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare double @modf(double noundef, double* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare float @modff(float noundef, float* noundef) [[NOT_READNONE]]
- // HAS_MAYTRAP: declare x86_fp80 @modfl(x86_fp80 noundef, x86_fp80* noundef) [[NOT_READNONE]]
+ // NO__ERRNO: declare double @modf(double noundef, ptr noundef) [[NOT_READNONE]]
+ // NO__ERRNO: declare float @modff(float noundef, ptr noundef) [[NOT_READNONE]]
+ // NO__ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare double @modf(double noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare float @modff(float noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_ERRNO: declare x86_fp80 @modfl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare double @modf(double noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare float @modff(float noundef, ptr noundef) [[NOT_READNONE]]
+ // HAS_MAYTRAP: declare x86_fp80 @modfl(x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
nan(c); nanf(c); nanl(c);
-// NO__ERRNO: declare double @nan(i8* noundef) [[READONLY:#[0-9]+]]
-// NO__ERRNO: declare float @nanf(i8* noundef) [[READONLY]]
-// NO__ERRNO: declare x86_fp80 @nanl(i8* noundef) [[READONLY]]
-// HAS_ERRNO: declare double @nan(i8* noundef) [[READONLY:#[0-9]+]]
-// HAS_ERRNO: declare float @nanf(i8* noundef) [[READONLY]]
-// HAS_ERRNO: declare x86_fp80 @nanl(i8* noundef) [[READONLY]]
-// HAS_MAYTRAP: declare double @nan(i8* noundef) [[READONLY:#[0-9]+]]
-// HAS_MAYTRAP: declare float @nanf(i8* noundef) [[READONLY]]
-// HAS_MAYTRAP: declare x86_fp80 @nanl(i8* noundef) [[READONLY]]
+// NO__ERRNO: declare double @nan(ptr noundef) [[READONLY:#[0-9]+]]
+// NO__ERRNO: declare float @nanf(ptr noundef) [[READONLY]]
+// NO__ERRNO: declare x86_fp80 @nanl(ptr noundef) [[READONLY]]
+// HAS_ERRNO: declare double @nan(ptr noundef) [[READONLY:#[0-9]+]]
+// HAS_ERRNO: declare float @nanf(ptr noundef) [[READONLY]]
+// HAS_ERRNO: declare x86_fp80 @nanl(ptr noundef) [[READONLY]]
+// HAS_MAYTRAP: declare double @nan(ptr noundef) [[READONLY:#[0-9]+]]
+// HAS_MAYTRAP: declare float @nanf(ptr noundef) [[READONLY]]
+// HAS_MAYTRAP: declare x86_fp80 @nanl(ptr noundef) [[READONLY]]
pow(f,f); powf(f,f); powl(f,f);
@@ -564,15 +564,15 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) {
remquo(f,f,i); remquof(f,f,i); remquol(f,f,i);
-// NO__ERRNO: declare double @remquo(double noundef, double noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare float @remquof(float noundef, float noundef, i32* noundef) [[NOT_READNONE]]
-// NO__ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare double @remquo(double noundef, double noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare float @remquof(float noundef, float noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_MAYTRAP: declare double @remquo(double noundef, double noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_MAYTRAP: declare float @remquof(float noundef, float noundef, i32* noundef) [[NOT_READNONE]]
-// HAS_MAYTRAP: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, i32* noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare double @remquo(double noundef, double noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare float @remquof(float noundef, float noundef, ptr noundef) [[NOT_READNONE]]
+// NO__ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare double @remquo(double noundef, double noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare float @remquof(float noundef, float noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_ERRNO: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_MAYTRAP: declare double @remquo(double noundef, double noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_MAYTRAP: declare float @remquof(float noundef, float noundef, ptr noundef) [[NOT_READNONE]]
+// HAS_MAYTRAP: declare x86_fp80 @remquol(x86_fp80 noundef, x86_fp80 noundef, ptr noundef) [[NOT_READNONE]]
rint(f); rintf(f); rintl(f);
diff --git a/clang/test/CodeGen/mult-alt-generic.c b/clang/test/CodeGen/mult-alt-generic.c
index 236acf27cd00f..581200b55da8a 100644
--- a/clang/test/CodeGen/mult-alt-generic.c
+++ b/clang/test/CodeGen/mult-alt-generic.c
@@ -1,14 +1,14 @@
-// RUN: %clang_cc1 -no-opaque-pointers -triple i686 %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64 %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple arm %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple mips %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple mipsel %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple powerpc %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple powerpc64 %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple s390x %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple sparc %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple sparcv9 %s -emit-llvm -o - | FileCheck %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple thumb %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple i686 %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple x86_64 %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple arm %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple mips %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple mipsel %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple powerpc %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple powerpc64 %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple s390x %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple sparc %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple sparcv9 %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -triple thumb %s -emit-llvm -o - | FileCheck %s
int mout0;
int min1;
@@ -17,7 +17,7 @@ int marray[2];
// CHECK: @single_m
void single_m(void)
{
- // CHECK: call void asm "foo $1,$0", "=*m,*m[[CLOBBERS:[a-zA-Z0-9@%{},~_$ ]*\"]](i32* elementtype(i32) {{[a-zA-Z0-9@%]+}}, i32* elementtype(i32) {{[a-zA-Z0-9@%]+}})
+ // CHECK: call void asm "foo $1,$0", "=*m,*m[[CLOBBERS:[a-zA-Z0-9@%{},~_$ ]*\"]](ptr elementtype(i32) {{[a-zA-Z0-9@%]+}}, ptr elementtype(i32) {{[a-zA-Z0-9@%]+}})
asm("foo %1,%0" : "=m" (mout0) : "m" (min1));
}
@@ -130,7 +130,7 @@ void single_X(void)
asm("foo %1,%0" : "=r" (out0) : "X" (min1));
// CHECK: call i32 asm "foo $1,$0", "=r,X[[CLOBBERS]](i32 1)
asm("foo %1,%0" : "=r" (out0) : "X" (1));
- // CHECK: call i32 asm "foo $1,$0", "=r,X[[CLOBBERS]](i32* getelementptr inbounds ([2 x i32], [2 x i32]* {{[a-zA-Z0-9@%]+}}, i{{32|64}} 0, i{{32|64}} 0))
+ // CHECK: call i32 asm "foo $1,$0", "=r,X[[CLOBBERS]](ptr {{[a-zA-Z0-9@%]+}})
asm("foo %1,%0" : "=r" (out0) : "X" (marray));
// CHECK: call i32 asm "foo $1,$0", "=r,X[[CLOBBERS]](double {{[0-9.eE+-]+}})
asm("foo %1,%0" : "=r" (out0) : "X" (1.0e+01));
@@ -143,14 +143,14 @@ void single_p(void)
{
register int out0 = 0;
// Constraint converted
diff erently on
diff erent platforms moved to platform-specific.
- // : call i32 asm "foo $1,$0", "=r,im[[CLOBBERS]](i32* getelementptr inbounds ([2 x i32], [2 x i32]* {{[a-zA-Z0-9@%]+}}, i{{32|64}} 0, i{{32|64}} 0))
+ // : call i32 asm "foo $1,$0", "=r,im[[CLOBBERS]](ptr getelementptr inbounds ([2 x i32], ptr {{[a-zA-Z0-9@%]+}}, i{{32|64}} 0, i{{32|64}} 0))
asm("foo %1,%0" : "=r" (out0) : "p" (marray));
}
// CHECK: @multi_m
void multi_m(void)
{
- // CHECK: call void asm "foo $1,$0", "=*m|r,m|r[[CLOBBERS]](i32* elementtype(i32) {{[a-zA-Z0-9@%]+}}, i32 {{[a-zA-Z0-9@%]+}})
+ // CHECK: call void asm "foo $1,$0", "=*m|r,m|r[[CLOBBERS]](ptr elementtype(i32) {{[a-zA-Z0-9@%]+}}, i32 {{[a-zA-Z0-9@%]+}})
asm("foo %1,%0" : "=m,r" (mout0) : "m,r" (min1));
}
@@ -263,7 +263,7 @@ void multi_X(void)
asm("foo %1,%0" : "=r,r" (out0) : "r,X" (min1));
// CHECK: call i32 asm "foo $1,$0", "=r|r,r|X[[CLOBBERS]](i32 1)
asm("foo %1,%0" : "=r,r" (out0) : "r,X" (1));
- // CHECK: call i32 asm "foo $1,$0", "=r|r,r|X[[CLOBBERS]](i32* getelementptr inbounds ([2 x i32], [2 x i32]* {{[a-zA-Z0-9@%]+}}, i{{32|64}} 0, i{{32|64}} 0))
+ // CHECK: call i32 asm "foo $1,$0", "=r|r,r|X[[CLOBBERS]](ptr {{[a-zA-Z0-9@%]+}})
asm("foo %1,%0" : "=r,r" (out0) : "r,X" (marray));
// CHECK: call i32 asm "foo $1,$0", "=r|r,r|X[[CLOBBERS]](double {{[0-9.eE+-]+}})
asm("foo %1,%0" : "=r,r" (out0) : "r,X" (1.0e+01));
@@ -276,6 +276,6 @@ void multi_p(void)
{
register int out0 = 0;
// Constraint converted
diff erently on
diff erent platforms moved to platform-specific.
- // : call i32 asm "foo $1,$0", "=r|r,r|im[[CLOBBERS]](i32* getelementptr inbounds ([2 x i32], [2 x i32]* {{[a-zA-Z0-9@%]+}}, {{i[0-9]*}} 0, {{i[0-9]*}} 0))
+ // : call i32 asm "foo $1,$0", "=r|r,r|im[[CLOBBERS]](ptr getelementptr inbounds ([2 x i32], ptr {{[a-zA-Z0-9@%]+}}, {{i[0-9]*}} 0, {{i[0-9]*}} 0))
asm("foo %1,%0" : "=r,r" (out0) : "r,p" (marray));
}
diff --git a/clang/test/CodeGenCUDA/device-stub.cu b/clang/test/CodeGenCUDA/device-stub.cu
index 8605df8a37702..d7a7b1bb9fe95 100644
--- a/clang/test/CodeGenCUDA/device-stub.cu
+++ b/clang/test/CodeGenCUDA/device-stub.cu
@@ -1,63 +1,63 @@
// RUN: echo -n "GPU binary would be here." > %t
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=8.0 -fcuda-include-gpubinary %t -o - \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s \
// RUN: --check-prefixes=ALL,LNX,NORDC,CUDA,CUDANORDC,CUDA-OLD
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=8.0 -fcuda-include-gpubinary %t \
// RUN: -o - -DNOGLOBALS \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s \
// RUN: -check-prefixes=NOGLOBALS,CUDANOGLOBALS
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=8.0 -fgpu-rdc -fcuda-include-gpubinary %t \
// RUN: -o - \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s \
// RUN: --check-prefixes=ALL,LNX,RDC,CUDA,CUDARDC,CUDA-OLD
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=8.0 -o - \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s -check-prefix=NOGPUBIN
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=9.2 -fcuda-include-gpubinary %t -o - \
// RUN: | FileCheck %s -allow-deprecated-dag-overlap \
// RUN: --check-prefixes=ALL,LNX,NORDC,CUDA,CUDANORDC,CUDA-NEW
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=9.2 -fcuda-include-gpubinary %t -o - -DNOGLOBALS \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s \
// RUN: --check-prefixes=NOGLOBALS,CUDANOGLOBALS
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=9.2 -fgpu-rdc -fcuda-include-gpubinary %t -o - \
// RUN: | FileCheck %s -allow-deprecated-dag-overlap \
// RUN: --check-prefixes=ALL,LNX,RDC,CUDA,CUDARDC,CUDA-NEW
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s -std=c++17 \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s -std=c++17 \
// RUN: -target-sdk-version=9.2 -fcuda-include-gpubinary %t -o - \
// RUN: | FileCheck %s -allow-deprecated-dag-overlap \
// RUN: --check-prefixes=ALL,LNX,NORDC,CUDA,CUDANORDC,CUDA-NEW,LNX_17,NORDC17
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s -std=c++17 \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s -std=c++17 \
// RUN: -target-sdk-version=9.2 -fgpu-rdc -fcuda-include-gpubinary %t -o - \
// RUN: | FileCheck %s -allow-deprecated-dag-overlap \
// RUN: --check-prefixes=ALL,LNX,RDC,CUDA,CUDARDC,CUDA-NEW,LNX_17,RDC17
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -target-sdk-version=9.2 -o - \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s -check-prefix=NOGPUBIN
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -fcuda-include-gpubinary %t -o - -x hip\
// RUN: | FileCheck -allow-deprecated-dag-overlap %s --check-prefixes=ALL,LNX,NORDC,HIP,HIPEF
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -fcuda-include-gpubinary %t -o - -DNOGLOBALS -x hip \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s -check-prefixes=NOGLOBALS,HIPNOGLOBALS
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s \
// RUN: -fgpu-rdc -fcuda-include-gpubinary %t -o - -x hip \
// RUN: | FileCheck -allow-deprecated-dag-overlap %s --check-prefixes=ALL,LNX,RDC,HIP,HIPEF
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-linux-gnu -emit-llvm %s -o - -x hip\
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -emit-llvm %s -o - -x hip\
// RUN: | FileCheck -allow-deprecated-dag-overlap %s -check-prefixes=ALL,LNX,NORDC,HIP,HIPNEF
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-pc-windows-msvc -aux-triple amdgcn -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-pc-windows-msvc -aux-triple amdgcn -emit-llvm %s \
// RUN: -fcuda-include-gpubinary %t -o - -x hip\
// RUN: | FileCheck -allow-deprecated-dag-overlap %s --check-prefixes=ALL,WIN
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-pc-windows-msvc -aux-triple amdgcn -emit-llvm %s \
+// RUN: %clang_cc1 -triple x86_64-pc-windows-msvc -aux-triple amdgcn -emit-llvm %s \
// RUN: -o - -x hip\
// RUN: | FileCheck -allow-deprecated-dag-overlap %s --check-prefixes=ALL,WIN,HIP,HIPNEF
@@ -166,26 +166,26 @@ __device__ void device_use() {
// CUDARDC-SAME: section "__nv_relfatbin", align 8
// * constant struct that wraps GPU binary
// ALL: @__[[PREFIX:cuda|hip]]_fatbin_wrapper = internal constant
-// LNX-SAME: { i32, i32, i8*, i8* }
+// LNX-SAME: { i32, i32, ptr, ptr }
// CUDA-SAME: { i32 1180844977, i32 1,
// HIP-SAME: { i32 1212764230, i32 1,
-// CUDA-SAME: i8* getelementptr inbounds ({{.*}}@[[FATBIN]], i64 0, i64 0),
-// HIPEF-SAME: i8* getelementptr inbounds ({{.*}}@[[FATBIN]], i64 0, i64 0),
-// HIPNEF-SAME: i8* @[[FATBIN]],
-// LNX-SAME: i8* null }
+// CUDA-SAME: ptr @[[FATBIN]],
+// HIPEF-SAME: ptr @[[FATBIN]],
+// HIPNEF-SAME: ptr @[[FATBIN]],
+// LNX-SAME: ptr null }
// CUDA-SAME: section ".nvFatBinSegment"
// HIP-SAME: section ".hipFatBinSegment"
// * variable to save GPU binary handle after initialization
-// CUDANORDC: @__[[PREFIX]]_gpubin_handle = internal global i8** null
-// HIPNEF: @__[[PREFIX]]_gpubin_handle = linkonce hidden global i8** null
+// CUDANORDC: @__[[PREFIX]]_gpubin_handle = internal global ptr null
+// HIPNEF: @__[[PREFIX]]_gpubin_handle = linkonce hidden global ptr null
// * constant unnamed string with NVModuleID
// CUDARDC: [[MODULE_ID_GLOBAL:@.*]] = private constant
// CUDARDC-SAME: c"[[MODULE_ID:.+]]\00", section "__nv_module_id", align 32
// * Make sure our constructor was added to global ctor list.
// LNX: @llvm.global_ctors = appending global {{.*}}@__[[PREFIX]]_module_ctor
// * Alias to global symbol containing the NVModuleID.
-// CUDARDC: @__fatbinwrap[[MODULE_ID]] ={{.*}} alias { i32, i32, i8*, i8* }
-// CUDARDC-SAME: { i32, i32, i8*, i8* }* @__[[PREFIX]]_fatbin_wrapper
+// CUDARDC: @__fatbinwrap[[MODULE_ID]] ={{.*}} alias { i32, i32, ptr, ptr }
+// CUDARDC-SAME: ptr @__[[PREFIX]]_fatbin_wrapper
// Test that we build the correct number of calls to cudaSetupArgument followed
// by a call to cudaLaunch.
@@ -225,13 +225,13 @@ void hostfunc(void) { kernelfunc<<<1, 1>>>(1, 1, 1); }
// Test that we've built a function to register kernels and global vars.
// ALL: define internal void @__[[PREFIX]]_register_globals
-// ALL: call{{.*}}[[PREFIX]]RegisterFunction(i8** %0, {{.*}}kernelfunc{{[^,]*}}, {{[^@]*}}@0
-// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(i8** %0, {{.*}}device_var{{[^,]*}}, {{[^@]*}}@1, {{.*}}i32 0, {{i32|i64}} 4, i32 0, i32 0
-// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(i8** %0, {{.*}}constant_var{{[^,]*}}, {{[^@]*}}@2, {{.*}}i32 0, {{i32|i64}} 4, i32 1, i32 0
-// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(i8** %0, {{.*}}ext_device_var_def{{[^,]*}}, {{[^@]*}}@3, {{.*}}i32 0, {{i32|i64}} 4, i32 0, i32 0
-// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(i8** %0, {{.*}}ext_constant_var_def{{[^,]*}}, {{[^@]*}}@4, {{.*}}i32 0, {{i32|i64}} 4, i32 1, i32 0
-// LNX_17-DAG: [[PREFIX]]RegisterVar(i8** %0, {{.*}}inline_var
-// LNX_17-NOT: [[PREFIX]]RegisterVar(i8** %0, {{.*}}inline_var2
+// ALL: call{{.*}}[[PREFIX]]RegisterFunction(ptr %0, {{.*}}kernelfunc{{[^,]*}}, {{[^@]*}}@0
+// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(ptr %0, {{.*}}device_var{{[^,]*}}, {{[^@]*}}@1, {{.*}}i32 0, {{i32|i64}} 4, i32 0, i32 0
+// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(ptr %0, {{.*}}constant_var{{[^,]*}}, {{[^@]*}}@2, {{.*}}i32 0, {{i32|i64}} 4, i32 1, i32 0
+// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(ptr %0, {{.*}}ext_device_var_def{{[^,]*}}, {{[^@]*}}@3, {{.*}}i32 0, {{i32|i64}} 4, i32 0, i32 0
+// ALL-DAG: call void {{.*}}[[PREFIX]]RegisterVar(ptr %0, {{.*}}ext_constant_var_def{{[^,]*}}, {{[^@]*}}@4, {{.*}}i32 0, {{i32|i64}} 4, i32 1, i32 0
+// LNX_17-DAG: [[PREFIX]]RegisterVar(ptr %0, {{.*}}inline_var
+// LNX_17-NOT: [[PREFIX]]RegisterVar(ptr %0, {{.*}}inline_var2
// ALL: ret void
// Test that we've built a constructor.
@@ -239,8 +239,8 @@ void hostfunc(void) { kernelfunc<<<1, 1>>>(1, 1, 1); }
// In separate mode it calls __[[PREFIX]]RegisterFatBinary(&__[[PREFIX]]_fatbin_wrapper)
// HIP only register fat binary once.
-// HIP: load i8**, i8*** @__hip_gpubin_handle
-// HIP-NEXT: icmp eq i8** {{.*}}, null
+// HIP: load ptr, ptr @__hip_gpubin_handle
+// HIP-NEXT: icmp eq ptr {{.*}}, null
// HIP-NEXT: br i1 {{.*}}, label %if, label %exit
// HIP: if:
// CUDANORDC: call{{.*}}[[PREFIX]]RegisterFatBinary{{.*}}__[[PREFIX]]_fatbin_wrapper
@@ -253,12 +253,12 @@ void hostfunc(void) { kernelfunc<<<1, 1>>>(1, 1, 1); }
// .. and then calls __[[PREFIX]]_register_globals
// HIP-NEXT: br label %exit
// HIP: exit:
-// HIP-NEXT: load i8**, i8*** @__hip_gpubin_handle
+// HIP-NEXT: load ptr, ptr @__hip_gpubin_handle
// CUDANORDC-NEXT: call void @__[[PREFIX]]_register_globals
// HIP-NEXT: call void @__[[PREFIX]]_register_globals
// * In separate mode we also register a destructor.
-// CUDANORDC-NEXT: call i32 @atexit(void ()* @__[[PREFIX]]_module_dtor)
-// HIP-NEXT: call i32 @atexit(void ()* @__[[PREFIX]]_module_dtor)
+// CUDANORDC-NEXT: call i32 @atexit(ptr @__[[PREFIX]]_module_dtor)
+// HIP-NEXT: call i32 @atexit(ptr @__[[PREFIX]]_module_dtor)
// With relocatable device code we call __[[PREFIX]]RegisterLinkedBinary%NVModuleID%
// CUDARDC: call{{.*}}__[[PREFIX]]RegisterLinkedBinary[[MODULE_ID]](
@@ -271,11 +271,11 @@ void hostfunc(void) { kernelfunc<<<1, 1>>>(1, 1, 1); }
// CUDANORDC: load{{.*}}__[[PREFIX]]_gpubin_handle
// HIP: load{{.*}}__[[PREFIX]]_gpubin_handle
// CUDANORDC-NEXT: call void @__[[PREFIX]]UnregisterFatBinary
-// HIP-NEXT: icmp ne i8** {{.*}}, null
+// HIP-NEXT: icmp ne ptr {{.*}}, null
// HIP-NEXT: br i1 {{.*}}, label %if, label %exit
// HIP: if:
// HIP-NEXT: call void @__[[PREFIX]]UnregisterFatBinary
-// HIP-NEXT: store i8** null, i8*** @__hip_gpubin_handle
+// HIP-NEXT: store ptr null, ptr @__hip_gpubin_handle
// HIP-NEXT: br label %exit
// HIP: exit:
diff --git a/clang/test/CodeGenCUDA/static-device-var-rdc.cu b/clang/test/CodeGenCUDA/static-device-var-rdc.cu
index a9cca4e9212ab..16ec413397235 100644
--- a/clang/test/CodeGenCUDA/static-device-var-rdc.cu
+++ b/clang/test/CodeGenCUDA/static-device-var-rdc.cu
@@ -1,19 +1,19 @@
// REQUIRES: x86-registered-target
// REQUIRES: amdgpu-registered-target
-// RUN: %clang_cc1 -no-opaque-pointers -triple amdgcn-amd-amdhsa -fcuda-is-device \
+// RUN: %clang_cc1 -triple amdgcn-amd-amdhsa -fcuda-is-device \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o %t.nocuid.dev -x hip %s
// RUN: cat %t.nocuid.dev | FileCheck -check-prefixes=DEV,INT-DEV %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-gnu-linux \
+// RUN: %clang_cc1 -triple x86_64-gnu-linux \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o %t.nocuid.host -x hip %s
// RUN: cat %t.nocuid.host | FileCheck -check-prefixes=HOST,INT-HOST %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple amdgcn-amd-amdhsa -fcuda-is-device -cuid=abc \
+// RUN: %clang_cc1 -triple amdgcn-amd-amdhsa -fcuda-is-device -cuid=abc \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s > %t.dev
// RUN: cat %t.dev | FileCheck -check-prefixes=DEV,EXT-DEV %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-gnu-linux -cuid=abc \
+// RUN: %clang_cc1 -triple x86_64-gnu-linux -cuid=abc \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s > %t.host
// RUN: cat %t.host | FileCheck -check-prefixes=HOST,EXT-HOST %s
@@ -25,25 +25,25 @@
// Negative tests.
-// RUN: %clang_cc1 -no-opaque-pointers -triple amdgcn-amd-amdhsa -fcuda-is-device \
+// RUN: %clang_cc1 -triple amdgcn-amd-amdhsa -fcuda-is-device \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s | FileCheck \
// RUN: -check-prefix=DEV-NEG %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-gnu-linux \
+// RUN: %clang_cc1 -triple x86_64-gnu-linux \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s | FileCheck \
// RUN: -check-prefix=HOST-NEG %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple amdgcn-amd-amdhsa -fcuda-is-device -cuid=abc \
+// RUN: %clang_cc1 -triple amdgcn-amd-amdhsa -fcuda-is-device -cuid=abc \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s > %t.dev
// RUN: cat %t.dev | FileCheck -check-prefix=DEV-NEG %s
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-gnu-linux -cuid=abc \
+// RUN: %clang_cc1 -triple x86_64-gnu-linux -cuid=abc \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - -x hip %s > %t.host
// RUN: cat %t.host | FileCheck -check-prefix=HOST-NEG %s
// Check postfix for CUDA.
-// RUN: %clang_cc1 -no-opaque-pointers -triple nvptx -fcuda-is-device -cuid=abc \
+// RUN: %clang_cc1 -triple nvptx -fcuda-is-device -cuid=abc \
// RUN: -std=c++11 -fgpu-rdc -emit-llvm -o - %s | FileCheck \
// RUN: -check-prefixes=CUDA %s
@@ -128,8 +128,8 @@ void foo() {
decltype(u) tmp;
}
-// HOST-DAG: __hipRegisterVar({{.*}}@_ZL1x {{.*}}@[[DEVNAMEX]]
-// HOST-DAG: __hipRegisterVar({{.*}}@_ZL1y {{.*}}@[[DEVNAMEY]]
+// HOST-DAG: __hipRegisterVar({{.*}}@_ZL1x, {{.*}}@[[DEVNAMEX]]
+// HOST-DAG: __hipRegisterVar({{.*}}@_ZL1y, {{.*}}@[[DEVNAMEY]]
// HOST-NEG-NOT: __hipRegisterVar({{.*}}@_ZL2x2
// HOST-NEG-NOT: __hipRegisterVar({{.*}}@_ZZ6kernelPiPPKiE1w
// HOST-NEG-NOT: __hipRegisterVar({{.*}}@_ZZ6devfunPPKiE1p
diff --git a/clang/test/CodeGenCXX/auto-var-init-stop-after.cpp b/clang/test/CodeGenCXX/auto-var-init-stop-after.cpp
index b87af1647fb7d..a782692d0127e 100644
--- a/clang/test/CodeGenCXX/auto-var-init-stop-after.cpp
+++ b/clang/test/CodeGenCXX/auto-var-init-stop-after.cpp
@@ -1,13 +1,13 @@
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=1 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-1-SCALAR
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=2 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-2-ARRAY
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=3 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-3-VLA
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=4 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-4-POINTER
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=5 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-5-BUILTIN
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=1 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-1-SCALAR
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=2 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-2-ARRAY
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=3 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-3-VLA
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=4 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-4-POINTER
-// RUN: %clang_cc1 -no-opaque-pointers -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=5 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-5-BUILTIN
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=1 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-1-SCALAR
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=2 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-2-ARRAY
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=3 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-3-VLA
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=4 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-4-POINTER
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=pattern -ftrivial-auto-var-init-stop-after=5 %s -emit-llvm -o - | FileCheck %s -check-prefix=PATTERN-STOP-AFTER-5-BUILTIN
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=1 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-1-SCALAR
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=2 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-2-ARRAY
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=3 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-3-VLA
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=4 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-4-POINTER
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ftrivial-auto-var-init=zero -ftrivial-auto-var-init-stop-after=5 %s -emit-llvm -o - | FileCheck %s -check-prefix=ZERO-STOP-AFTER-5-BUILTIN
#define ARRLEN 10
@@ -27,32 +27,29 @@ int foo(unsigned n) {
void *p;
// builtin
p = __builtin_alloca(sizeof(unsigned long long) * n);
- // PATTERN-STOP-AFTER-1-SCALAR: store i64 -6148914691236517206, i64* %a, align 8
- // PATTERN-STOP-AFTER-1-SCALAR-NOT: call void @llvm.memset.p0i8.i64(i8* align 16 %0, i8 -86, i64 80, i1 false)
- // PATTERN-STOP-AFTER-2-ARRAY: %0 = bitcast [10 x %struct.S]* %arr to i8*
- // PATTERN-STOP-AFTER-2-ARRAY-NEXT: call void @llvm.memset.p0i8.i64(i8* align 16 %0, i8 -86, i64 80, i1 false)
+ // PATTERN-STOP-AFTER-1-SCALAR: store i64 -6148914691236517206, ptr %a, align 8
+ // PATTERN-STOP-AFTER-1-SCALAR-NOT: call void @llvm.memset.p0.i64(ptr align 16 %0, i8 -86, i64 80, i1 false)
+ // PATTERN-STOP-AFTER-2-ARRAY: call void @llvm.memset.p0.i64(ptr align 16 %arr, i8 -86, i64 80, i1 false)
// PATTERN-STOP-AFTER-2-ARRAY-NOT: vla-init.loop:
// PATTERN-STOP-AFTER-3-VLA: vla-init.loop:
- // PATTERN-STOP-AFTER-3-VLA-NEXT: %vla.cur = phi i8* [ %vla.begin, %vla-setup.loop ], [ %vla.next, %vla-init.loop ]
- // PATTERN-STOP-AFTER-3-VLA-NEXT-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* align 8 %vla.cur, i8* align 4 bitcast ({ i32, i8, [3 x i8] }* @__const._Z3fooj.vla to i8*), i64 8, i1 false)
- // PATTERN-STOP-AFTER-3-VLA-NOT: store i8* inttoptr (i64 -6148914691236517206 to i8*), i8** %p, align 8
- // PATTERN-STOP-AFTER-4-POINTER: store i8* inttoptr (i64 -6148914691236517206 to i8*), i8** %p, align 8
- // PATTERN-STOP-AFTER-4-POINTER-NOT: call void @llvm.memset.p0i8.i64(i8* align 16 %6, i8 -86, i64 %mul, i1 false)
- // PATTERN-STOP-AFTER-5-BUILTIN: call void @llvm.memset.p0i8.i64(i8* align 16 %6, i8 -86, i64 %mul, i1 false)
+ // PATTERN-STOP-AFTER-3-VLA-NEXT: %vla.cur = phi ptr [ %vla, %vla-setup.loop ], [ %vla.next, %vla-init.loop ]
+ // PATTERN-STOP-AFTER-3-VLA-NEXT-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 8 %vla.cur, ptr align 4 @__const._Z3fooj.vla, i64 8, i1 false)
+ // PATTERN-STOP-AFTER-3-VLA-NOT: store ptr inttoptr (i64 -6148914691236517206 to ptr), ptr %p, align 8
+ // PATTERN-STOP-AFTER-4-POINTER: store ptr inttoptr (i64 -6148914691236517206 to ptr), ptr %p, align 8
+ // PATTERN-STOP-AFTER-4-POINTER-NOT: call void @llvm.memset.p0.i64(ptr align 16 %5, i8 -86, i64 %mul, i1 false)
+ // PATTERN-STOP-AFTER-5-BUILTIN: call void @llvm.memset.p0.i64(ptr align 16 %5, i8 -86, i64 %mul, i1 false)
// PATTERN-STOP-AFTER-5-BUILTIN-MESSAGES: warning: -ftrivial-auto-var-init-stop-after=5 has been enabled to limit the number of times ftrivial-auto-var-init=pattern gets applied.
- // ZERO-STOP-AFTER-1-SCALAR: store i64 0, i64* %a, align 8
- // ZERO-STOP-AFTER-1-SCALAR-NOT: call void @llvm.memset.p0i8.i64(i8* align 16 %0, i8 0, i64 80, i1 false)
- // ZERO-STOP-AFTER-2-ARRAY: %0 = bitcast [10 x %struct.S]* %arr to i8*
- // ZERO-STOP-AFTER-2-ARRAY-NEXT: call void @llvm.memset.p0i8.i64(i8* align 16 %0, i8 0, i64 80, i1 false)
- // ZERO-STOP-AFTER-2-ARRAY-NOT: %call void @llvm.memset.p0i8.i64(i8* align 16 %5, i8 0, i64 %4, i1 false)
- // ZERO-STOP-AFTER-3-VLA: %5 = bitcast %struct.S* %vla to i8*
- // ZERO-STOP-AFTER-3-VLA-NEXT: call void @llvm.memset.p0i8.i64(i8* align 16 %5, i8 0, i64 %4, i1 false)
- // ZERO-STOP-AFTER-3-VLA-NOT: store i8* null, i8** %p, align 8
- // ZERO-STOP-AFTER-4-POINTER: store i8* null, i8** %p, align 8
- // ZERO-STOP-AFTER-4-POINTER-NOT: call void @llvm.memset.p0i8.i64(i8* align 16 %7, i8 0, i64 %mul, i1 false)
- // ZERO-STOP-AFTER-5-BUILTIN: %7 = alloca i8, i64 %mul, align 16
- // ZERO-STOP-AFTER-5-BUILTIN-NEXT: call void @llvm.memset.p0i8.i64(i8* align 16 %7, i8 0, i64 %mul, i1 false)
+ // ZERO-STOP-AFTER-1-SCALAR: store i64 0, ptr %a, align 8
+ // ZERO-STOP-AFTER-1-SCALAR-NOT: call void @llvm.memset.p0.i64(ptr align 16 %arr, i8 0, i64 80, i1 false)
+ // ZERO-STOP-AFTER-2-ARRAY: call void @llvm.memset.p0.i64(ptr align 16 %arr, i8 0, i64 80, i1 false)
+ // ZERO-STOP-AFTER-2-ARRAY-NOT: %call void @llvm.memset.p0.i64(ptr align 16 %3, i8 0, i64 %2, i1 false)
+ // ZERO-STOP-AFTER-3-VLA: call void @llvm.memset.p0.i64(ptr align 16 %vla, i8 0, i64 %3, i1 false)
+ // ZERO-STOP-AFTER-3-VLA-NOT: store ptr null, ptr %p, align 8
+ // ZERO-STOP-AFTER-4-POINTER: store ptr null, ptr %p, align 8
+ // ZERO-STOP-AFTER-4-POINTER-NOT: call void @llvm.memset.p0.i64(ptr align 16 %4, i8 0, i64 %mul, i1 false)
+ // ZERO-STOP-AFTER-5-BUILTIN: %5 = alloca i8, i64 %mul, align 16
+ // ZERO-STOP-AFTER-5-BUILTIN-NEXT: call void @llvm.memset.p0.i64(ptr align 16 %5, i8 0, i64 %mul, i1 false)
// ZERO-STOP-AFTER-5-BUILTIN-MESSAGES: warnings: -ftrivial-auto-var-init-stop-after=5 has been enabled to limit the number of times ftrivial-auto-var-init=zero gets applied.
return 0;
}
More information about the cfe-commits
mailing list