diff --git a/clang/include/clang/Basic/Builtins.def b/clang/include/clang/Basic/Builtins.def index fe00a2f69922a..c0e045865e2c2 100644 --- a/clang/include/clang/Basic/Builtins.def +++ b/clang/include/clang/Basic/Builtins.def @@ -143,7 +143,6 @@ BUILTIN(__builtin_frexp , "ddi*" , "Fn") BUILTIN(__builtin_frexpf, "ffi*" , "Fn") BUILTIN(__builtin_frexpl, "LdLdi*", "Fn") BUILTIN(__builtin_frexpf128, "LLdLLdi*", "Fn") -BUILTIN(__builtin_frexpf16, "hhi*" , "Fn") BUILTIN(__builtin_huge_val, "d", "ncE") BUILTIN(__builtin_huge_valf, "f", "ncE") BUILTIN(__builtin_huge_vall, "Ld", "ncE") diff --git a/clang/lib/CodeGen/CGBuiltin.cpp b/clang/lib/CodeGen/CGBuiltin.cpp index bb59d2ea90f55..3a0588ad752b8 100644 --- a/clang/lib/CodeGen/CGBuiltin.cpp +++ b/clang/lib/CodeGen/CGBuiltin.cpp @@ -652,24 +652,6 @@ emitMaybeConstrainedFPToIntRoundBuiltin(CodeGenFunction &CGF, const CallExpr *E, } } -static Value *emitFrexpBuiltin(CodeGenFunction &CGF, const CallExpr *E, - llvm::Intrinsic::ID IntrinsicID) { - llvm::Value *Src0 = CGF.EmitScalarExpr(E->getArg(0)); - llvm::Value *Src1 = CGF.EmitScalarExpr(E->getArg(1)); - - QualType IntPtrTy = E->getArg(1)->getType()->getPointeeType(); - llvm::Type *IntTy = CGF.ConvertType(IntPtrTy); - llvm::Function *F = - CGF.CGM.getIntrinsic(IntrinsicID, {Src0->getType(), IntTy}); - llvm::Value *Call = CGF.Builder.CreateCall(F, Src0); - - llvm::Value *Exp = CGF.Builder.CreateExtractValue(Call, 1); - LValue LV = CGF.MakeNaturalAlignAddrLValue(Src1, IntPtrTy); - CGF.EmitStoreOfScalar(Exp, LV); - - return CGF.Builder.CreateExtractValue(Call, 0); -} - /// EmitFAbs - Emit a call to @llvm.fabs(). static Value *EmitFAbs(CodeGenFunction &CGF, Value *V) { Function *F = CGF.CGM.getIntrinsic(Intrinsic::fabs, V->getType()); @@ -3080,12 +3062,6 @@ RValue CodeGenFunction::EmitBuiltinExpr(const GlobalDecl GD, unsigned BuiltinID, { Src0->getType(), Src1->getType() }); return RValue::get(Builder.CreateCall(F, { Src0, Src1 })); } - case Builtin::BI__builtin_frexp: - case Builtin::BI__builtin_frexpf: - case Builtin::BI__builtin_frexpl: - case Builtin::BI__builtin_frexpf128: - case Builtin::BI__builtin_frexpf16: - return RValue::get(emitFrexpBuiltin(*this, E, Intrinsic::frexp)); case Builtin::BI__builtin_isgreater: case Builtin::BI__builtin_isgreaterequal: case Builtin::BI__builtin_isless: diff --git a/clang/lib/Headers/__clang_hip_math.h b/clang/lib/Headers/__clang_hip_math.h index a914496cb7b14..c67959673f936 100644 --- a/clang/lib/Headers/__clang_hip_math.h +++ b/clang/lib/Headers/__clang_hip_math.h @@ -257,7 +257,8 @@ float fmodf(float __x, float __y) { return __ocml_fmod_f32(__x, __y); } __DEVICE__ float frexpf(float __x, int *__nptr) { - return __builtin_frexpf(__x, __nptr); + *__nptr = __builtin_amdgcn_frexp_expf(__x); + return __builtin_amdgcn_frexp_mantf(__x); } __DEVICE__ @@ -805,7 +806,8 @@ double fmod(double __x, double __y) { return __ocml_fmod_f64(__x, __y); } __DEVICE__ double frexp(double __x, int *__nptr) { - return __builtin_frexp(__x, __nptr); + *__nptr = __builtin_amdgcn_frexp_exp(__x); + return __builtin_amdgcn_frexp_mant(__x); } __DEVICE__ diff --git a/clang/test/CodeGen/aix-builtin-mapping.c b/clang/test/CodeGen/aix-builtin-mapping.c index a79218c6f1d8b..98fcfd4a3a6fc 100644 --- a/clang/test/CodeGen/aix-builtin-mapping.c +++ b/clang/test/CodeGen/aix-builtin-mapping.c @@ -18,5 +18,5 @@ int main() } // CHECK: %call = call double @modf(double noundef 1.000000e+00, ptr noundef %DummyLongDouble) #3 -// CHECK: %{{.+}} = call { double, i32 } @llvm.frexp.f64.i32(double 0.000000e+00) +// CHECK: %call1 = call double @frexp(double noundef 0.000000e+00, ptr noundef %DummyInt) #3 // CHECK: %{{.+}} = call double @llvm.ldexp.f64.i32(double 1.000000e+00, i32 1) diff --git a/clang/test/CodeGen/builtin-attributes.c b/clang/test/CodeGen/builtin-attributes.c index ec184e3a7b048..fd1e107a41d58 100644 --- a/clang/test/CodeGen/builtin-attributes.c +++ b/clang/test/CodeGen/builtin-attributes.c @@ -4,10 +4,6 @@ int printf(const char *, ...); void exit(int); -float frexpf(float, int*); -double frexp(double, int*); -long double frexpl(long double, int*); - // CHECK: declare i32 @printf(ptr noundef, ...) void f0() { printf("a\n"); @@ -53,9 +49,9 @@ char* f2(char* a, char* b) { // CHECK: ret int f3(double x) { int e; - frexp(x, &e); - frexpf(x, &e); - frexpl(x, &e); + __builtin_frexp(x, &e); + __builtin_frexpf(x, &e); + __builtin_frexpl(x, &e); __builtin_modf(x, &e); __builtin_modff(x, &e); __builtin_modfl(x, &e); diff --git a/clang/test/CodeGen/math-builtins-long.c b/clang/test/CodeGen/math-builtins-long.c index f3c328dcbfcd7..e32fc99480ead 100644 --- a/clang/test/CodeGen/math-builtins-long.c +++ b/clang/test/CodeGen/math-builtins-long.c @@ -34,10 +34,10 @@ void foo(long double f, long double *l, int *i, const char *c) { // PPCF128: call fp128 @llvm.fabs.f128(fp128 %{{.+}}) __builtin_fabsl(f); - // F80: call { x86_fp80, i32 } @llvm.frexp.f80.i32(x86_fp80 %{{.+}}) - // PPC: call { ppc_fp128, i32 } @llvm.frexp.ppcf128.i32(ppc_fp128 %{{.+}}) - // X86F128: call { fp128, i32 } @llvm.frexp.f128.i32(fp128 %{{.+}}) - // PPCF128: call { fp128, i32 } @llvm.frexp.f128.i32(fp128 %{{.+}}) + // F80: call x86_fp80 @frexpl(x86_fp80 noundef %{{.+}}, ptr noundef %{{.+}}) + // PPC: call ppc_fp128 @frexpl(ppc_fp128 noundef %{{.+}}, ptr noundef %{{.+}}) + // X86F128: call fp128 @frexpl(fp128 noundef %{{.+}}, ptr noundef %{{.+}}) + // PPCF128: call fp128 @frexpf128(fp128 noundef %{{.+}}, ptr noundef %{{.+}}) __builtin_frexpl(f,i); // F80: store x86_fp80 0xK7FFF8000000000000000, ptr diff --git a/clang/test/CodeGen/math-builtins.c b/clang/test/CodeGen/math-builtins.c index 257da44b5888a..6233501937920 100644 --- a/clang/test/CodeGen/math-builtins.c +++ b/clang/test/CodeGen/math-builtins.c @@ -12,30 +12,6 @@ void foo(double *d, float f, float *fp, long double *l, int *i, const char *c) { // NO__ERRNO: frem float // NO__ERRNO: frem x86_fp80 // NO__ERRNO: frem fp128 - -// NO__ERRNO: [[FREXP_F64:%.+]] = call { double, i32 } @llvm.frexp.f64.i32(double %{{.+}}) -// NO__ERRNO-NEXT: [[FREXP_F64_1:%.+]] = extractvalue { double, i32 } [[FREXP_F64]], 1 -// NO__ERRNO-NEXT: store i32 [[FREXP_F64_1]], ptr %{{.+}}, align 4 -// NO__ERRNO-NEXT: [[FREXP_F64_0:%.+]] = extractvalue { double, i32 } [[FREXP_F64]], 0 - -// NO__ERRNO: [[FREXP_F32:%.+]] = call { float, i32 } @llvm.frexp.f32.i32(float %{{.+}}) -// NO__ERRNO-NEXT: [[FREXP_F32_1:%.+]] = extractvalue { float, i32 } [[FREXP_F32]], 1 -// NO__ERRNO-NEXT: store i32 [[FREXP_F32_1]], ptr %{{.+}}, align 4 -// NO__ERRNO-NEXT: [[FREXP_F32_0:%.+]] = extractvalue { float, i32 } [[FREXP_F32]], 0 - - -// NO__ERRNO: [[FREXP_F80:%.+]] = call { x86_fp80, i32 } @llvm.frexp.f80.i32(x86_fp80 %{{.+}}) -// NO__ERRNO-NEXT: [[FREXP_F80_1:%.+]] = extractvalue { x86_fp80, i32 } [[FREXP_F80]], 1 -// NO__ERRNO-NEXT: store i32 [[FREXP_F80_1]], ptr %{{.+}}, align 4 -// NO__ERRNO-NEXT: [[FREXP_F80_0:%.+]] = extractvalue { x86_fp80, i32 } [[FREXP_F80]], 0 - - -// NO__ERRNO: [[FREXP_F128:%.+]] = call { fp128, i32 } @llvm.frexp.f128.i32(fp128 %{{.+}}) -// NO__ERRNO-NEXT: [[FREXP_F128_1:%.+]] = extractvalue { fp128, i32 } [[FREXP_F128]], 1 -// NO__ERRNO-NEXT: store i32 [[FREXP_F128_1]], ptr %{{.+}}, align 4 -// NO__ERRNO-NEXT: [[FREXP_F128_0:%.+]] = extractvalue { fp128, i32 } [[FREXP_F128]], 0 - - // HAS_ERRNO: declare double @fmod(double noundef, double noundef) [[NOT_READNONE:#[0-9]+]] // HAS_ERRNO: declare float @fmodf(float noundef, float noundef) [[NOT_READNONE]] // HAS_ERRNO: declare x86_fp80 @fmodl(x86_fp80 noundef, x86_fp80 noundef) [[NOT_READNONE]] @@ -76,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, i32 } @llvm.frexp.f64.i32(double) [[READNONE_INTRINSIC]] -// NO__ERRNO: declare { float, i32 } @llvm.frexp.f32.i32(float) [[READNONE_INTRINSIC]] -// NO__ERRNO: declare { x86_fp80, i32 } @llvm.frexp.f80.i32(x86_fp80) [[READNONE_INTRINSIC]] -// NO__ERRNO: declare { fp128, i32 } @llvm.frexp.f128.i32(fp128) [[READNONE_INTRINSIC]] -// HAS_ERRNO: declare { double, i32 } @llvm.frexp.f64.i32(double) [[READNONE_INTRINSIC]] -// HAS_ERRNO: declare { float, i32 } @llvm.frexp.f32.i32(float) [[READNONE_INTRINSIC]] -// HAS_ERRNO: declare { x86_fp80, i32 } @llvm.frexp.f80.i32(x86_fp80) [[READNONE_INTRINSIC]] -// HAS_ERRNO: declare { fp128, i32 } @llvm.frexp.f128.i32(fp128) [[READNONE_INTRINSIC]] +// 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(); @@ -112,7 +88,7 @@ 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, ptr noundef) [[NOT_READNONE:#[0-9]+]] +// 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]] diff --git a/clang/test/CodeGenOpenCL/builtins-generic-amdgcn.cl b/clang/test/CodeGenOpenCL/builtins-generic-amdgcn.cl index b093fcbf7d989..aa8f13310b5fd 100644 --- a/clang/test/CodeGenOpenCL/builtins-generic-amdgcn.cl +++ b/clang/test/CodeGenOpenCL/builtins-generic-amdgcn.cl @@ -39,33 +39,3 @@ float test_builtin_ldexpf(float v, int e) { double test_builtin_ldexp(double v, int e) { return __builtin_ldexp(v, e); } - -// CHECK-LABEL: @test_builtin_frexpf16( -// CHECK: [[VAL:%.+]] = tail call { half, i32 } @llvm.frexp.f16.i32(half %v) -// CHECK: [[EXTRACT_1:%.+]] = extractvalue { half, i32 } [[VAL]], 1 -// CHECK: store i32 [[EXTRACT_1]], ptr addrspace(5) -// CHECK: [[EXTRACT_0:%.+]] = extractvalue { half, i32 } [[VAL]], 0 -// CHECK: ret half [[EXTRACT_0]] -half test_builtin_frexpf16(half v, int* e) { - return __builtin_frexpf16(v, e); -} - -// CHECK-LABEL: @test_builtin_frexpf( -// CHECK: [[VAL:%.+]] = tail call { float, i32 } @llvm.frexp.f32.i32(float %v) -// CHECK: [[EXTRACT_1:%.+]] = extractvalue { float, i32 } [[VAL]], 1 -// CHECK: store i32 [[EXTRACT_1]], ptr addrspace(5) -// CHECK: [[EXTRACT_0:%.+]] = extractvalue { float, i32 } [[VAL]], 0 -// CHECK: ret float [[EXTRACT_0]] -float test_builtin_frexpf(float v, int* e) { - return __builtin_frexpf(v, e); -} - -// CHECK-LABEL: @test_builtin_frexp( -// CHECK: [[VAL:%.+]] = tail call { double, i32 } @llvm.frexp.f64.i32(double %v) -// CHECK: [[EXTRACT_1:%.+]] = extractvalue { double, i32 } [[VAL]], 1 -// CHECK: store i32 [[EXTRACT_1]], ptr addrspace(5) -// CHECK: [[EXTRACT_0:%.+]] = extractvalue { double, i32 } [[VAL]], 0 -// CHECK: ret double [[EXTRACT_0]] -double test_builtin_frexp(double v, int* e) { - return __builtin_frexp(v, e); -} diff --git a/clang/test/Headers/__clang_hip_math.hip b/clang/test/Headers/__clang_hip_math.hip index 8e5201a89ceca..c4a848c9bc621 100644 --- a/clang/test/Headers/__clang_hip_math.hip +++ b/clang/test/Headers/__clang_hip_math.hip @@ -1061,25 +1061,37 @@ extern "C" __device__ double test_fmod(double x, double y) { return fmod(x, y); } -// CHECK-LABEL: @test_frexpf( -// CHECK-NEXT: entry: -// CHECK-NEXT: [[TMP0:%.*]] = tail call { float, i32 } @llvm.frexp.f32.i32(float [[X:%.*]]) -// CHECK-NEXT: [[TMP1:%.*]] = extractvalue { float, i32 } [[TMP0]], 1 -// CHECK-NEXT: store i32 [[TMP1]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12:![0-9]+]] -// CHECK-NEXT: [[TMP2:%.*]] = extractvalue { float, i32 } [[TMP0]], 0 -// CHECK-NEXT: ret float [[TMP2]] +// DEFAULT-LABEL: @test_frexpf( +// DEFAULT-NEXT: entry: +// DEFAULT-NEXT: [[TMP0:%.*]] = tail call i32 @llvm.amdgcn.frexp.exp.i32.f32(float [[X:%.*]]) +// DEFAULT-NEXT: store i32 [[TMP0]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12:![0-9]+]] +// DEFAULT-NEXT: [[TMP1:%.*]] = tail call contract float @llvm.amdgcn.frexp.mant.f32(float [[X]]) +// DEFAULT-NEXT: ret float [[TMP1]] +// +// FINITEONLY-LABEL: @test_frexpf( +// FINITEONLY-NEXT: entry: +// FINITEONLY-NEXT: [[TMP0:%.*]] = tail call i32 @llvm.amdgcn.frexp.exp.i32.f32(float [[X:%.*]]) +// FINITEONLY-NEXT: store i32 [[TMP0]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12:![0-9]+]] +// FINITEONLY-NEXT: [[TMP1:%.*]] = tail call nnan ninf contract float @llvm.amdgcn.frexp.mant.f32(float [[X]]) +// FINITEONLY-NEXT: ret float [[TMP1]] // extern "C" __device__ float test_frexpf(float x, int* y) { return frexpf(x, y); } -// CHECK-LABEL: @test_frexp( -// CHECK-NEXT: entry: -// CHECK-NEXT: [[TMP0:%.*]] = tail call { double, i32 } @llvm.frexp.f64.i32(double [[X:%.*]]) -// CHECK-NEXT: [[TMP1:%.*]] = extractvalue { double, i32 } [[TMP0]], 1 -// CHECK-NEXT: store i32 [[TMP1]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12]] -// CHECK-NEXT: [[TMP2:%.*]] = extractvalue { double, i32 } [[TMP0]], 0 -// CHECK-NEXT: ret double [[TMP2]] +// DEFAULT-LABEL: @test_frexp( +// DEFAULT-NEXT: entry: +// DEFAULT-NEXT: [[TMP0:%.*]] = tail call i32 @llvm.amdgcn.frexp.exp.i32.f64(double [[X:%.*]]) +// DEFAULT-NEXT: store i32 [[TMP0]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12]] +// DEFAULT-NEXT: [[TMP1:%.*]] = tail call contract double @llvm.amdgcn.frexp.mant.f64(double [[X]]) +// DEFAULT-NEXT: ret double [[TMP1]] +// +// FINITEONLY-LABEL: @test_frexp( +// FINITEONLY-NEXT: entry: +// FINITEONLY-NEXT: [[TMP0:%.*]] = tail call i32 @llvm.amdgcn.frexp.exp.i32.f64(double [[X:%.*]]) +// FINITEONLY-NEXT: store i32 [[TMP0]], ptr [[Y:%.*]], align 4, !tbaa [[TBAA12]] +// FINITEONLY-NEXT: [[TMP1:%.*]] = tail call nnan ninf contract double @llvm.amdgcn.frexp.mant.f64(double [[X]]) +// FINITEONLY-NEXT: ret double [[TMP1]] // extern "C" __device__ double test_frexp(double x, int* y) { return frexp(x, y);