[clang] ffe12e7 - clang: Handle MatrixType in hasFloatingRepresentation
Matt Arsenault via cfe-commits
cfe-commits at lists.llvm.org
Tue Mar 14 22:13:26 PDT 2023
Author: Matt Arsenault
Date: 2023-03-15T01:13:08-04:00
New Revision: ffe12e765cb9ba77dc30ccec89fb163234b98541
URL: https://github.com/llvm/llvm-project/commit/ffe12e765cb9ba77dc30ccec89fb163234b98541
DIFF: https://github.com/llvm/llvm-project/commit/ffe12e765cb9ba77dc30ccec89fb163234b98541.diff
LOG: clang: Handle MatrixType in hasFloatingRepresentation
Allows applying nofpclass to matrix arguments.
Added:
Modified:
clang/lib/AST/Type.cpp
clang/test/CodeGen/matrix-type-operators-fast-math.c
clang/test/CodeGen/nofpclass.c
Removed:
################################################################################
diff --git a/clang/lib/AST/Type.cpp b/clang/lib/AST/Type.cpp
index 96e611968ca18..71a098c254244 100644
--- a/clang/lib/AST/Type.cpp
+++ b/clang/lib/AST/Type.cpp
@@ -2154,8 +2154,9 @@ bool Type::isFloatingType() const {
bool Type::hasFloatingRepresentation() const {
if (const auto *VT = dyn_cast<VectorType>(CanonicalType))
return VT->getElementType()->isFloatingType();
- else
- return isFloatingType();
+ if (const auto *MT = dyn_cast<MatrixType>(CanonicalType))
+ return MT->getElementType()->isFloatingType();
+ return isFloatingType();
}
bool Type::isRealFloatingType() const {
diff --git a/clang/test/CodeGen/matrix-type-operators-fast-math.c b/clang/test/CodeGen/matrix-type-operators-fast-math.c
index f71faa63f4e70..724b282830286 100644
--- a/clang/test/CodeGen/matrix-type-operators-fast-math.c
+++ b/clang/test/CodeGen/matrix-type-operators-fast-math.c
@@ -8,7 +8,7 @@ typedef unsigned long long ullx4x2_t __attribute__((matrix_type(4, 2)));
// Floating point matrix/scalar additions.
void add_matrix_matrix_double(dx5x5_t a, dx5x5_t b, dx5x5_t c) {
- // CHECK-LABEL: define{{.*}} void @add_matrix_matrix_double(<25 x double> noundef %a, <25 x double> noundef %b, <25 x double> noundef %c)
+ // CHECK-LABEL: define{{.*}} void @add_matrix_matrix_double(<25 x double> noundef nofpclass(nan inf) %a, <25 x double> noundef nofpclass(nan inf) %b, <25 x double> noundef nofpclass(nan inf) %c)
// CHECK: [[B:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[C:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[RES:%.*]] = fadd reassoc nnan ninf nsz arcp afn <25 x double> [[B]], [[C]]
@@ -18,7 +18,7 @@ void add_matrix_matrix_double(dx5x5_t a, dx5x5_t b, dx5x5_t c) {
}
void add_compound_assign_matrix_double(dx5x5_t a, dx5x5_t b) {
- // CHECK-LABEL: define{{.*}} void @add_compound_assign_matrix_double(<25 x double> noundef %a, <25 x double> noundef %b)
+ // CHECK-LABEL: define{{.*}} void @add_compound_assign_matrix_double(<25 x double> noundef nofpclass(nan inf) %a, <25 x double> noundef nofpclass(nan inf) %b)
// CHECK: [[B:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[A:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[RES:%.*]] = fadd reassoc nnan ninf nsz arcp afn <25 x double> [[A]], [[B]]
@@ -28,7 +28,7 @@ void add_compound_assign_matrix_double(dx5x5_t a, dx5x5_t b) {
}
void subtract_compound_assign_matrix_double(dx5x5_t a, dx5x5_t b) {
- // CHECK-LABEL: define{{.*}} void @subtract_compound_assign_matrix_double(<25 x double> noundef %a, <25 x double> noundef %b)
+ // CHECK-LABEL: define{{.*}} void @subtract_compound_assign_matrix_double(<25 x double> noundef nofpclass(nan inf) %a, <25 x double> noundef nofpclass(nan inf) %b)
// CHECK: [[B:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[A:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[RES:%.*]] = fsub reassoc nnan ninf nsz arcp afn <25 x double> [[A]], [[B]]
@@ -38,7 +38,7 @@ void subtract_compound_assign_matrix_double(dx5x5_t a, dx5x5_t b) {
}
void add_matrix_scalar_double_float(dx5x5_t a, float vf) {
- // CHECK-LABEL: define{{.*}} void @add_matrix_scalar_double_float(<25 x double> noundef %a, float noundef nofpclass(nan inf) %vf)
+ // CHECK-LABEL: define{{.*}} void @add_matrix_scalar_double_float(<25 x double> noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %vf)
// CHECK: [[MATRIX:%.*]] = load <25 x double>, ptr {{.*}}, align 8
// CHECK-NEXT: [[SCALAR:%.*]] = load float, ptr %vf.addr, align 4
// CHECK-NEXT: [[SCALAR_EXT:%.*]] = fpext float [[SCALAR]] to double
@@ -51,7 +51,7 @@ void add_matrix_scalar_double_float(dx5x5_t a, float vf) {
}
void add_compound_matrix_scalar_double_float(dx5x5_t a, float vf) {
- // CHECK-LABEL: define{{.*}} void @add_compound_matrix_scalar_double_float(<25 x double> noundef %a, float noundef nofpclass(nan inf) %vf)
+ // CHECK-LABEL: define{{.*}} void @add_compound_matrix_scalar_double_float(<25 x double> noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %vf)
// CHECK: [[SCALAR:%.*]] = load float, ptr %vf.addr, align 4
// CHECK-NEXT: [[SCALAR_EXT:%.*]] = fpext float [[SCALAR]] to double
// CHECK-NEXT: [[MATRIX:%.*]] = load <25 x double>, ptr {{.*}}, align 8
@@ -64,7 +64,7 @@ void add_compound_matrix_scalar_double_float(dx5x5_t a, float vf) {
}
void subtract_compound_matrix_scalar_double_float(dx5x5_t a, float vf) {
- // CHECK-LABEL: define{{.*}} void @subtract_compound_matrix_scalar_double_float(<25 x double> noundef %a, float noundef nofpclass(nan inf) %vf)
+ // CHECK-LABEL: define{{.*}} void @subtract_compound_matrix_scalar_double_float(<25 x double> noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %vf)
// CHECK: [[SCALAR:%.*]] = load float, ptr %vf.addr, align 4
// CHECK-NEXT: [[SCALAR_EXT:%.*]] = fpext float [[SCALAR]] to double
// CHECK-NEXT: [[MATRIX:%.*]] = load <25 x double>, ptr {{.*}}, align 8
diff --git a/clang/test/CodeGen/nofpclass.c b/clang/test/CodeGen/nofpclass.c
index 5f28ef877e83d..7fea8192cda75 100644
--- a/clang/test/CodeGen/nofpclass.c
+++ b/clang/test/CodeGen/nofpclass.c
@@ -1396,40 +1396,40 @@ __m256d call_m256d(__m256d x) {
}
// CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local <25 x double> @call_matrix
-// CFINITEONLY-SAME: (<25 x double> noundef [[X:%.*]]) #[[ATTR6:[0-9]+]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <25 x double> @call_matrix
+// CFINITEONLY-SAME: (<25 x double> noundef nofpclass(nan inf) [[X:%.*]]) #[[ATTR6:[0-9]+]] {
// CFINITEONLY-NEXT: entry:
// CFINITEONLY-NEXT: [[X_ADDR:%.*]] = alloca [25 x double], align 8
// CFINITEONLY-NEXT: store <25 x double> [[X]], ptr [[X_ADDR]], align 8
// CFINITEONLY-NEXT: [[TMP0:%.*]] = load <25 x double>, ptr [[X_ADDR]], align 8
-// CFINITEONLY-NEXT: [[CALL:%.*]] = call nnan ninf <25 x double> @extern_matrix(<25 x double> noundef [[TMP0]])
+// CFINITEONLY-NEXT: [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) <25 x double> @extern_matrix(<25 x double> noundef nofpclass(nan inf) [[TMP0]])
// CFINITEONLY-NEXT: ret <25 x double> [[CALL]]
//
// CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local <25 x double> @call_matrix
-// CLFINITEONLY-SAME: (<25 x double> noundef [[X:%.*]]) local_unnamed_addr #[[ATTR9:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <25 x double> @call_matrix
+// CLFINITEONLY-SAME: (<25 x double> noundef nofpclass(nan inf) [[X:%.*]]) local_unnamed_addr #[[ATTR9:[0-9]+]] {
// CLFINITEONLY-NEXT: entry:
-// CLFINITEONLY-NEXT: [[CALL:%.*]] = tail call nnan ninf <25 x double> @extern_matrix(<25 x double> noundef [[X]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT: [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) <25 x double> @extern_matrix(<25 x double> noundef nofpclass(nan inf) [[X]]) #[[ATTR10]]
// CLFINITEONLY-NEXT: ret <25 x double> [[CALL]]
//
// NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local <25 x double> @call_matrix
-// NONANS-SAME: (<25 x double> noundef [[X:%.*]]) #[[ATTR6:[0-9]+]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) <25 x double> @call_matrix
+// NONANS-SAME: (<25 x double> noundef nofpclass(nan) [[X:%.*]]) #[[ATTR6:[0-9]+]] {
// NONANS-NEXT: entry:
// NONANS-NEXT: [[X_ADDR:%.*]] = alloca [25 x double], align 8
// NONANS-NEXT: store <25 x double> [[X]], ptr [[X_ADDR]], align 8
// NONANS-NEXT: [[TMP0:%.*]] = load <25 x double>, ptr [[X_ADDR]], align 8
-// NONANS-NEXT: [[CALL:%.*]] = call nnan <25 x double> @extern_matrix(<25 x double> noundef [[TMP0]])
+// NONANS-NEXT: [[CALL:%.*]] = call nnan nofpclass(nan) <25 x double> @extern_matrix(<25 x double> noundef nofpclass(nan) [[TMP0]])
// NONANS-NEXT: ret <25 x double> [[CALL]]
//
// NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local <25 x double> @call_matrix
-// NOINFS-SAME: (<25 x double> noundef [[X:%.*]]) #[[ATTR6:[0-9]+]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) <25 x double> @call_matrix
+// NOINFS-SAME: (<25 x double> noundef nofpclass(inf) [[X:%.*]]) #[[ATTR6:[0-9]+]] {
// NOINFS-NEXT: entry:
// NOINFS-NEXT: [[X_ADDR:%.*]] = alloca [25 x double], align 8
// NOINFS-NEXT: store <25 x double> [[X]], ptr [[X_ADDR]], align 8
// NOINFS-NEXT: [[TMP0:%.*]] = load <25 x double>, ptr [[X_ADDR]], align 8
-// NOINFS-NEXT: [[CALL:%.*]] = call ninf <25 x double> @extern_matrix(<25 x double> noundef [[TMP0]])
+// NOINFS-NEXT: [[CALL:%.*]] = call ninf nofpclass(inf) <25 x double> @extern_matrix(<25 x double> noundef nofpclass(inf) [[TMP0]])
// NOINFS-NEXT: ret <25 x double> [[CALL]]
//
dx5x5_t call_matrix(dx5x5_t x) {
More information about the cfe-commits
mailing list