Skip to content

Commit 2aeefcf

Browse files
paperchalicepaperchalice
authored
[clang][CodeGen] Remove "unsafe-fp-math" attribute support (#162779)
These global flags block furthur improvements for clang, users should always use fast-math flags see also https://discourse.llvm.org/t/rfc-honor-pragmas-with-ffp-contract-fast/80797 Remove them incrementally, this is the clang part.
1 parent 187a8e3 commit 2aeefcf

File tree

8 files changed

+15
-43
lines changed

8 files changed

+15
-43
lines changed

‎clang/lib/CodeGen/CGCall.cpp‎

Lines changed: 0 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -2012,13 +2012,6 @@ static void getTrivialDefaultFunctionAttributes(
20122012
FuncAttrs.addAttribute("no-infs-fp-math", "true");
20132013
if (LangOpts.NoHonorNaNs)
20142014
FuncAttrs.addAttribute("no-nans-fp-math", "true");
2015-
if (LangOpts.AllowFPReassoc && LangOpts.AllowRecip &&
2016-
LangOpts.NoSignedZero && LangOpts.ApproxFunc &&
2017-
(LangOpts.getDefaultFPContractMode() ==
2018-
LangOptions::FPModeKind::FPM_Fast ||
2019-
LangOpts.getDefaultFPContractMode() ==
2020-
LangOptions::FPModeKind::FPM_FastHonorPragmas))
2021-
FuncAttrs.addAttribute("unsafe-fp-math", "true");
20222015
if (CodeGenOpts.SoftFloat)
20232016
FuncAttrs.addAttribute("use-soft-float", "true");
20242017
FuncAttrs.addAttribute("stack-protector-buffer-size",

‎clang/lib/CodeGen/CGCall.h‎

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -410,10 +410,10 @@ class ReturnValueSlot {
410410
/// This is useful for adding attrs to bitcode modules that you want to link
411411
/// with but don't control, such as CUDA's libdevice. When linking with such
412412
/// a bitcode library, you might want to set e.g. its functions'
413-
/// "unsafe-fp-math" attribute to match the attr of the functions you're
413+
/// "denormal-fp-math" attribute to match the attr of the functions you're
414414
/// codegen'ing. Otherwise, LLVM will interpret the bitcode module's lack of
415-
/// unsafe-fp-math attrs as tantamount to unsafe-fp-math=false, and then LLVM
416-
/// will propagate unsafe-fp-math=false up to every transitive caller of a
415+
/// denormal-fp-math attrs as tantamount to denormal-fp-math=ieee, and then LLVM
416+
/// will propagate denormal-fp-math=ieee up to every transitive caller of a
417417
/// function in the bitcode library!
418418
///
419419
/// With the exception of fast-math attrs, this will only make the attributes

‎clang/lib/CodeGen/CodeGenFunction.cpp‎

Lines changed: 0 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -183,11 +183,6 @@ void CodeGenFunction::CGFPOptionsRAII::ConstructorHelper(FPOptions FPFeatures) {
183183
mergeFnAttrValue("no-infs-fp-math", FPFeatures.getNoHonorInfs());
184184
mergeFnAttrValue("no-nans-fp-math", FPFeatures.getNoHonorNaNs());
185185
mergeFnAttrValue("no-signed-zeros-fp-math", FPFeatures.getNoSignedZero());
186-
mergeFnAttrValue(
187-
"unsafe-fp-math",
188-
FPFeatures.getAllowFPReassociate() && FPFeatures.getAllowReciprocal() &&
189-
FPFeatures.getAllowApproxFunc() && FPFeatures.getNoSignedZero() &&
190-
FPFeatures.allowFPContractAcrossStatement());
191186
}
192187

193188
CodeGenFunction::CGFPOptionsRAII::~CGFPOptionsRAII() {

‎clang/test/CodeGen/backend-unsupported-error.ll‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -21,7 +21,7 @@ entry:
2121
ret i32 %call, !dbg !15
2222
}
2323

24-
attributes #0 = { nounwind noinline "disable-tail-calls"="false" "less-precise-fpmad"="false" "frame-pointer"="all" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "stack-protector-buffer-size"="8" "unsafe-fp-math"="false" "use-soft-float"="false" }
24+
attributes #0 = { nounwind noinline "disable-tail-calls"="false" "less-precise-fpmad"="false" "frame-pointer"="all" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "stack-protector-buffer-size"="8" "use-soft-float"="false" }
2525

2626
!llvm.dbg.cu = !{!0}
2727
!llvm.module.flags = !{!9, !10}

‎clang/test/CodeGen/fp-function-attrs.cpp‎

Lines changed: 4 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -36,7 +36,7 @@ float test_reassociate_off_pragma(float a, float b, float c) {
3636
return tmp;
3737
}
3838

39-
// CHECK: define{{.*}} float @_Z27test_reassociate_off_pragmafff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c) [[NO_UNSAFE_ATTRS:#[0-9]+]]
39+
// CHECK: define{{.*}} float @_Z27test_reassociate_off_pragmafff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c)
4040
// CHECK: fadd nnan ninf nsz arcp contract afn float {{%.+}}, {{%.+}}
4141
// CHECK: fadd fast float {{%.+}}, {{%.+}}
4242

@@ -49,10 +49,9 @@ float test_contract_on_pragma(float a, float b, float c) {
4949
return tmp;
5050
}
5151

52-
// CHECK: define{{.*}} float @_Z23test_contract_on_pragmafff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c) [[NO_UNSAFE_ATTRS:#[0-9]+]]
52+
// CHECK: define{{.*}} float @_Z23test_contract_on_pragmafff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c)
5353
// CHECK: fmul fast float {{%.+}}, {{%.+}}
5454
// CHECK: fadd reassoc nnan ninf nsz arcp afn float {{%.+}}, {{%.+}}
5555

56-
// CHECK: attributes [[FAST_ATTRS]] = { {{.*}}"no-infs-fp-math"="true" {{.*}}"no-nans-fp-math"="true" "no-signed-zeros-fp-math"="true" {{.*}}"unsafe-fp-math"="true"{{.*}} }
57-
// CHECK: attributes [[PRECISE_ATTRS]] = { {{.*}}"no-infs-fp-math"="false" {{.*}}"no-nans-fp-math"="false" "no-signed-zeros-fp-math"="false" {{.*}}"unsafe-fp-math"="false"{{.*}} }
58-
// CHECK: attributes [[NO_UNSAFE_ATTRS]] = { {{.*}}"no-infs-fp-math"="true" {{.*}}"no-nans-fp-math"="true" "no-signed-zeros-fp-math"="true" {{.*}}"unsafe-fp-math"="false"{{.*}} }
56+
// CHECK: attributes [[FAST_ATTRS]] = { {{.*}}"no-infs-fp-math"="true" {{.*}}"no-nans-fp-math"="true" "no-signed-zeros-fp-math"="true"{{.*}} }
57+
// CHECK: attributes [[PRECISE_ATTRS]] = { {{.*}}"no-infs-fp-math"="false" {{.*}}"no-nans-fp-math"="false" "no-signed-zeros-fp-math"="false"{{.*}} }

‎clang/test/CodeGen/func-attr.c‎

Lines changed: 4 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -1,18 +1,18 @@
11
// RUN: %clang_cc1 -triple x86_64-linux-gnu -ffast-math \
22
// RUN: -ffp-contract=fast -emit-llvm -o - %s | \
3-
// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE,FINITEONLY
3+
// RUN: FileCheck %s --check-prefixes=CHECK,FINITEONLY
44

55
// RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
66
// RUN: -ffp-contract=fast -emit-llvm -o - %s | \
7-
// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE,NOFINITEONLY
7+
// RUN: FileCheck %s --check-prefixes=CHECK,NOFINITEONLY
88

99
// RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
1010
// RUN: -ffp-contract=on -emit-llvm -o - %s | \
11-
// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE,NOFINITEONLY
11+
// RUN: FileCheck %s --check-prefixes=CHECK,NOFINITEONLY
1212

1313
// RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
1414
// RUN: -ffp-contract=off -emit-llvm -o - %s | \
15-
// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE,NOFINITEONLY
15+
// RUN: FileCheck %s --check-prefixes=CHECK,NOFINITEONLY
1616

1717
float foo(float a, float b) {
1818
return a+b;
@@ -24,6 +24,4 @@ float foo(float a, float b) {
2424
// CHECK: attributes [[ATTRS]] = {
2525
// CHECK-SAME: "no-signed-zeros-fp-math"="true"
2626
// CHECK-SAME: "no-trapping-math"="true"
27-
// CHECK-UNSAFE-SAME: "unsafe-fp-math"="true"
28-
// CHECK-NOUNSAFE-NOT: "unsafe-fp-math"="true"
2927
// CHECK-SAME: }

‎clang/test/CodeGenCUDA/propagate-attributes.cu‎

Lines changed: 3 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -12,18 +12,17 @@
1212

1313
// RUN: %clang_cc1 -x cuda %s -emit-llvm -mlink-builtin-bitcode %t.bc -o - \
1414
// RUN: -fcuda-is-device -triple nvptx-unknown-unknown \
15-
// RUN: | FileCheck %s --check-prefix=CHECK --check-prefix=NOFTZ --check-prefix=NOFAST
15+
// RUN: | FileCheck %s --check-prefix=CHECK --check-prefix=NOFTZ
1616

1717
// RUN: %clang_cc1 -x cuda %s -emit-llvm -mlink-builtin-bitcode %t.bc \
1818
// RUN: -fdenormal-fp-math-f32=preserve-sign -o - \
1919
// RUN: -fcuda-is-device -triple nvptx-unknown-unknown \
20-
// RUN: | FileCheck %s --check-prefix=CHECK --check-prefix=FTZ \
21-
// RUN: --check-prefix=NOFAST
20+
// RUN: | FileCheck %s --check-prefix=CHECK --check-prefix=FTZ
2221

2322
// RUN: %clang_cc1 -x cuda %s -emit-llvm -mlink-builtin-bitcode %t.bc \
2423
// RUN: -fdenormal-fp-math-f32=preserve-sign -o - \
2524
// RUN: -fcuda-is-device -funsafe-math-optimizations -triple nvptx-unknown-unknown \
26-
// RUN: | FileCheck %s --check-prefix=CHECK --check-prefix=FAST
25+
// RUN: | FileCheck %s --check-prefix=CHECK
2726

2827
#ifndef LIB
2928
#include "Inputs/cuda.h"
@@ -65,9 +64,6 @@ __global__ void kernel() { lib_fn(); }
6564

6665
// CHECK-SAME: "no-trapping-math"="true"
6766

68-
// FAST-SAME: "unsafe-fp-math"="true"
69-
// NOFAST-NOT: "unsafe-fp-math"="true"
70-
7167
// Check the attribute list for lib_fn.
7268
// CHECK: attributes [[fattr]] = {
7369

@@ -81,6 +77,3 @@ __global__ void kernel() { lib_fn(); }
8177
// NOFTZ-NOT: "denormal-fp-math-f32"
8278

8379
// CHECK-SAME: "no-trapping-math"="true"
84-
85-
// FAST-SAME: "unsafe-fp-math"="true"
86-
// NOFAST-NOT: "unsafe-fp-math"="true"

‎clang/test/CodeGenOpenCL/relaxed-fpmath.cl‎

Lines changed: 0 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -33,37 +33,31 @@ float spscalardiv(float a, float b) {
3333
// NORMAL-NOT: "no-infs-fp-math"
3434
// NORMAL-NOT: "no-nans-fp-math"
3535
// NORMAL-NOT: "no-signed-zeros-fp-math"
36-
// NORMAL-NOT: "unsafe-fp-math"
3736

3837
// FAST: "less-precise-fpmad"="true"
3938
// FAST: "no-infs-fp-math"="true"
4039
// FAST: "no-nans-fp-math"="true"
4140
// FAST: "no-signed-zeros-fp-math"="true"
42-
// FAST: "unsafe-fp-math"="true"
4341

4442
// FINITE-NOT: "less-precise-fpmad"
4543
// FINITE: "no-infs-fp-math"="true"
4644
// FINITE: "no-nans-fp-math"="true"
4745
// FINITE-NOT: "no-signed-zeros-fp-math"
48-
// FINITE-NOT: "unsafe-fp-math"
4946

5047
// UNSAFE: "less-precise-fpmad"="true"
5148
// UNSAFE-NOT: "no-infs-fp-math"
5249
// UNSAFE-NOT: "no-nans-fp-math"
5350
// UNSAFE: "no-signed-zeros-fp-math"="true"
54-
// UNSAFE: "unsafe-fp-math"="true"
5551

5652
// MAD: "less-precise-fpmad"="true"
5753
// MAD-NOT: "no-infs-fp-math"
5854
// MAD-NOT: "no-nans-fp-math"
5955
// MAD-NOT: "no-signed-zeros-fp-math"
60-
// MAD-NOT: "unsafe-fp-math"
6156

6257
// NOSIGNED-NOT: "less-precise-fpmad"
6358
// NOSIGNED-NOT: "no-infs-fp-math"
6459
// NOSIGNED-NOT: "no-nans-fp-math"
6560
// NOSIGNED: "no-signed-zeros-fp-math"="true"
66-
// NOSIGNED-NOT: "unsafe-fp-math"
6761

6862
#else
6963
// Undefine this to avoid putting it in the PCH.

0 commit comments

Comments
 (0)