clang: Emit nofpclass(nan inf) for -ffinite-math-only

[arsenm]

Summary

Set this on any source level floating-point type argument,
return value, call return or outgoing parameter which is lowered
to a valid IR type for the attribute. Currently this isn't
applied to emitted intrinsics since those don't go through
ABI code.

[jcranmer-intel]

clang/lib/CodeGen/CGCall.cpp

2212

Clang doesn't have support for -f[no-]signaling-nans yet, but the gcc documentation for the option states:

Compile code assuming that IEEE signaling NaNs may generate user-visible traps during floating-point operations. Setting this option disables optimizations that may change the number of exceptions visible with signaling NaNs. This option implies -ftrapping-math.

This strikes me as saying that sNaNs are treated as qNaN (akin to the nsz fast-math flag) rather than saying that it's UB to have sNaN as a value, thus, I don't think it makes sense for -fno-signaling-nans to translate into a nofpclass attribute.

[arsenm]

Drop todo

[nikic]

(Sounds reasonable conceptually, but I'm not familiar enough with clang codegen for the details.)

[jcranmer-intel]

I'm generally okay with the approach of this patch. I'm not sufficiently well-versed in the clang codegen side of things to know if this covers all of the bases, and I'd appreciate someone who is familiar with that side of things to approve this patch.

[tra]

I'm in the same boat as @nikic -- the patch looks reasonable in principle, but I'm not familiar with the nuanaces of FP codegen.

[efriedma]

Looks fine from a codegen perspective, assuming these are the semantics we want for -ffinite-math-only.

clang/lib/CodeGen/CGCall.cpp

3053

What cases does this cover that aren't already covered by the code in CodeGenModule::ConstructAttributeList?

[arsenm]

Drop unnecessary part

[efriedma]

LGTM

[arsenm]

dd8181055494d8f973dfaaba282bb1ae6eab4d40

clang/lib/CodeGen/CGCall.cpp

3053

Nothing apparently

clang/lib/CodeGen/CGCall.cpp

if (ArgNo < FD->getNumParams()) {
     const ParmVarDecl *Param = FD->getParamDecl(ArgNo);
     if (Param && Param->hasAttr<MaybeUndefAttr>())
       return true;
   }
 
   return false;
 }
 
+/// Test if it's legal to apply nofpclass for the given parameter type and it's
+/// lowered IR type.
+static bool canApplyNoFPClass(const ABIArgInfo &AI, QualType ParamType,
+                              bool IsReturn) {
+  // Should only apply to FP types in the source, not ABI promoted.
+  if (!ParamType->hasFloatingRepresentation())
+    return false;
+
+  // The promoted-to IR type also needs to support nofpclass.
+  llvm::Type *IRTy = AI.getCoerceToType();
+  if (llvm::AttributeFuncs::isNoFPClassCompatibleType(IRTy))
+    return true;
+
+  if (llvm::StructType *ST = dyn_cast<llvm::StructType>(IRTy)) {
+    return !IsReturn && AI.getCanBeFlattened() &&
+           llvm::all_of(ST->elements(), [](llvm::Type *Ty) {
+             return llvm::AttributeFuncs::isNoFPClassCompatibleType(Ty);
+           });
+  }
+
+  return false;
+}
+
+/// Return the nofpclass mask that can be applied to floating-point parameters.
+static llvm::FPClassTest getNoFPClassTestMask(const LangOptions &LangOpts) {
+  // TODO: Handle -fno-signaling-nans
+  llvm::FPClassTest Mask = llvm::fcNone;
+  if (LangOpts.NoHonorInfs)
+    Mask |= llvm::fcInf;
+  if (LangOpts.NoHonorNaNs)
+    Mask |= llvm::fcNan;
+  return Mask;
+}
+
 /// Construct the IR attribute list of a function or call.
 ///
 /// When adding an attribute, please consider where it should be handled:
 ///
 ///   - getDefaultFunctionAttributes is for attributes that are essentially
 ///     part of the global target configuration (but perhaps can be
 ///     overridden on a per-function basis).  Adding attributes there
 ///     will cause them to also be set in frontends that build on Clang's
case ABIArgInfo::Extend:
     if (RetAI.isSignExt())
       RetAttrs.addAttribute(llvm::Attribute::SExt);
     else
       RetAttrs.addAttribute(llvm::Attribute::ZExt);
     [[fallthrough]];
   case ABIArgInfo::Direct:
     if (RetAI.getInReg())
       RetAttrs.addAttribute(llvm::Attribute::InReg);
+
+    if (canApplyNoFPClass(RetAI, RetTy, true))
+      RetAttrs.addNoFPClassAttr(getNoFPClassTestMask(getLangOpts()));
+
     break;
   case ABIArgInfo::Ignore:
     break;
 
   case ABIArgInfo::InAlloca:
   case ABIArgInfo::Indirect: {
     // inalloca and sret disable readnone and readonly
     AddPotentialArgAccess();
case ABIArgInfo::Extend:
         Attrs.addAttribute(llvm::Attribute::ZExt);
       [[fallthrough]];
     case ABIArgInfo::Direct:
       if (ArgNo == 0 && FI.isChainCall())
         Attrs.addAttribute(llvm::Attribute::Nest);
       else if (AI.getInReg())
         Attrs.addAttribute(llvm::Attribute::InReg);
       Attrs.addStackAlignmentAttr(llvm::MaybeAlign(AI.getDirectAlign()));
-      break;
 
+      if (canApplyNoFPClass(AI, ParamType, false))
+        Attrs.addNoFPClassAttr(getNoFPClassTestMask(getLangOpts()));
+      break;
     case ABIArgInfo::Indirect: {
       if (AI.getInReg())
         Attrs.addAttribute(llvm::Attribute::InReg);
 
       if (AI.getIndirectByVal())
         Attrs.addByValAttr(getTypes().ConvertTypeForMem(ParamType));
 
       auto *Decl = ParamType->getAsRecordDecl();
case ABIArgInfo::Direct: {
           }
         }
 
         // Set 'noalias' if an argument type has the `restrict` qualifier.
         if (Arg->getType().isRestrictQualified())
           AI->addAttr(llvm::Attribute::NoAlias);
       }
 
+      if (canApplyNoFPClass(ArgI, Arg->getType(), false)) {
+        llvm::FPClassTest Mask = getNoFPClassTestMask(getLangOpts());
+        AI->addAttrs(
+            llvm::AttrBuilder(getLLVMContext()).addNoFPClassAttr(Mask));
+      }
+
       // Prepare the argument value. If we have the trivial case, handle it
       // with no muss and fuss.
       if (!isa<llvm::StructType>(ArgI.getCoerceToType()) &&
           ArgI.getCoerceToType() == ConvertType(Ty) &&
           ArgI.getDirectOffset() == 0) {
         assert(NumIRArgs == 1);
 
         // LLVM expects swifterror parameters to be used in very restricted

clang/test/CodeGen/complex-math.c

float _Complex div_float_rc(float a, float _Complex b) {
   // X86-LABEL: @div_float_rc(
   // X86-NOT: fdiv
   // X86: call {{.*}} @__divsc3(
   // X86: ret
 
   // SPIR: call spir_func {{.*}} @__divsc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_float_rc(float noundef %a, [2 x float] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_float_rc(float noundef nofpclass(nan inf) %a, [2 x float] noundef nofpclass(nan inf) %b.coerce)
   // A = a
   // B = 0
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast float
   // BD = 0
   // ACpBD = AC
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast float
float _Complex div_float_cc(float _Complex a, float _Complex b) {
   // X86-LABEL: @div_float_cc(
   // X86-NOT: fdiv
   // X86: call {{.*}} @__divsc3(
   // X86: ret
 
   // SPIR: call spir_func {{.*}} @__divsc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_float_cc([2 x float] noundef %a.coerce, [2 x float] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_float_cc([2 x float] noundef nofpclass(nan inf) %a.coerce, [2 x float] noundef nofpclass(nan inf) %b.coerce)
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast float
   // AARCH64-FASTMATH: [[BD:%.*]] = fmul fast float
   // AARCH64-FASTMATH: [[ACpBD:%.*]] = fadd fast float
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast float
   // AARCH64-FASTMATH: [[DD:%.*]] = fmul fast float
   // AARCH64-FASTMATH: [[CCpDD:%.*]] = fadd fast float
double _Complex div_double_rc(double a, double _Complex b) {
   // X86-LABEL: @div_double_rc(
   // X86-NOT: fdiv
   // X86: call {{.*}} @__divdc3(
   // X86: ret
 
   // SPIR: call spir_func {{.*}} @__divdc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_double_rc(double noundef %a, [2 x double] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_double_rc(double noundef nofpclass(nan inf) %a, [2 x double] noundef nofpclass(nan inf) %b.coerce)
   // A = a
   // B = 0
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast double
   // BD = 0
   // ACpBD = AC
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast double
double _Complex div_double_cc(double _Complex a, double _Complex b) {
   // X86-LABEL: @div_double_cc(
   // X86-NOT: fdiv
   // X86: call {{.*}} @__divdc3(
   // X86: ret
 
   // SPIR: call spir_func {{.*}} @__divdc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_double_cc([2 x double] noundef %a.coerce, [2 x double] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_double_cc([2 x double] noundef nofpclass(nan inf) %a.coerce, [2 x double] noundef nofpclass(nan inf) %b.coerce)
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast double
   // AARCH64-FASTMATH: [[BD:%.*]] = fmul fast double
   // AARCH64-FASTMATH: [[ACpBD:%.*]] = fadd fast double
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast double
   // AARCH64-FASTMATH: [[DD:%.*]] = fmul fast double
   // AARCH64-FASTMATH: [[CCpDD:%.*]] = fadd fast double
long double _Complex div_long_double_rc(long double a, long double _Complex b) {
   // X86: ret
   // PPC-LABEL: @div_long_double_rc(
   // PPC-NOT: fdiv
   // PPC: call {{.*}} @__divtc3(
   // PPC: ret
   // SPIR: call spir_func {{.*}} @__divdc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_long_double_rc(fp128 noundef %a, [2 x fp128] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_long_double_rc(fp128 noundef nofpclass(nan inf) %a, [2 x fp128] noundef nofpclass(nan inf) %b.coerce)
   // A = a
   // B = 0
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast fp128
   // BD = 0
   // ACpBD = AC
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast fp128
long double _Complex div_long_double_cc(long double _Complex a, long double _Complex b) {
   // X86: ret
   // PPC-LABEL: @div_long_double_cc(
   // PPC-NOT: fdiv
   // PPC: call {{.*}} @__divtc3(
   // PPC: ret
   // SPIR: call spir_func {{.*}} @__divdc3(
 
   // a / b = (A+iB) / (C+iD) = ((AC+BD)/(CC+DD)) + i((BC-AD)/(CC+DD))
-  // AARCH64-FASTMATH-LABEL: @div_long_double_cc([2 x fp128] noundef %a.coerce, [2 x fp128] noundef %b.coerce)
+  // AARCH64-FASTMATH-LABEL: @div_long_double_cc([2 x fp128] noundef nofpclass(nan inf) %a.coerce, [2 x fp128] noundef nofpclass(nan inf) %b.coerce)
   //
   // AARCH64-FASTMATH: [[AC:%.*]] = fmul fast fp128
   // AARCH64-FASTMATH: [[BD:%.*]] = fmul fast fp128
   // AARCH64-FASTMATH: [[ACpBD:%.*]] = fadd fast fp128
   //
   // AARCH64-FASTMATH: [[CC:%.*]] = fmul fast fp128
   // AARCH64-FASTMATH: [[DD:%.*]] = fmul fast fp128
   // AARCH64-FASTMATH: [[CCpDD:%.*]] = fadd fast fp128

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

 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffast-math -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffast-math -ffp-contract=fast-honor-pragmas -emit-llvm -o - %s | FileCheck %s
 
 float test_default(float a, float b, float c) {
   float tmp = a;
   tmp += b;
   tmp += c;
   return tmp;
 }
 
-// CHECK: define{{.*}} float @_Z12test_defaultfff(float noundef %a, float noundef %b, float noundef %c) [[FAST_ATTRS:#[0-9]+]]
+// CHECK: define{{.*}} float @_Z12test_defaultfff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c) [[FAST_ATTRS:#[0-9]+]]
 // CHECK: fadd fast float {{%.+}}, {{%.+}}
 // CHECK: fadd fast float {{%.+}}, {{%.+}}
 
 float test_precise_on_pragma(float a, float b, float c) {
   float tmp = a;
   {
     #pragma float_control(precise, on)
     tmp += b;
   }
   tmp += c;
   return tmp;
 }
 
-// CHECK: define{{.*}} float @_Z22test_precise_on_pragmafff(float noundef %a, float noundef %b, float noundef %c) [[PRECISE_ATTRS:#[0-9]+]]
+// CHECK: define{{.*}} float @_Z22test_precise_on_pragmafff(float noundef nofpclass(nan inf) %a, float noundef nofpclass(nan inf) %b, float noundef nofpclass(nan inf) %c) [[PRECISE_ATTRS:#[0-9]+]]
 // CHECK: fadd float {{%.+}}, {{%.+}}
 // CHECK: fadd fast float {{%.+}}, {{%.+}}
 
 float test_reassociate_off_pragma(float a, float b, float c) {
   float tmp = a;
   {
     #pragma clang fp reassociate(off)
     tmp += b;
   }
   tmp += c;
   return tmp;
 }
 
-// CHECK: define{{.*}} float @_Z27test_reassociate_off_pragmafff(float noundef %a, float noundef %b, float noundef %c) [[NO_UNSAFE_ATTRS:#[0-9]+]]
+// 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]+]]
 // CHECK: fadd nnan ninf nsz arcp contract afn float {{%.+}}, {{%.+}}
 // CHECK: fadd fast float {{%.+}}, {{%.+}}
 
 float test_contract_on_pragma(float a, float b, float c) {
   float tmp = a * b;
   {
     #pragma clang fp contract(on)
     tmp += c;
   }
   return tmp;
 }
 
-// CHECK: define{{.*}} float @_Z23test_contract_on_pragmafff(float noundef %a, float noundef %b, float noundef %c) [[NO_UNSAFE_ATTRS:#[0-9]+]]
+// 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]+]]
 // CHECK: fmul fast float {{%.+}}, {{%.+}}
 // CHECK: fadd reassoc nnan ninf nsz arcp afn float {{%.+}}, {{%.+}}
 
 // CHECK: attributes [[FAST_ATTRS]] = { {{.*}}"no-infs-fp-math"="true" {{.*}}"no-nans-fp-math"="true" "no-signed-zeros-fp-math"="true" {{.*}}"unsafe-fp-math"="true"{{.*}} }
 // CHECK: attributes [[PRECISE_ATTRS]] = { {{.*}}"no-infs-fp-math"="false" {{.*}}"no-nans-fp-math"="false" "no-signed-zeros-fp-math"="false" {{.*}}"unsafe-fp-math"="false"{{.*}} }
 // 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"{{.*}} }

clang/test/CodeGen/fp-options-to-fast-math-flags.c

 
 float test(float a) {
   return a + fn(a);
 }
 
 // CHECK-PRECISE: [[CALL_RES:%.+]] = call float @fn(float noundef {{%.+}})
 // CHECK-PRECISE: {{%.+}} = fadd float {{%.+}}, [[CALL_RES]]
 
-// CHECK-NO-NANS: [[CALL_RES:%.+]] = call nnan float @fn(float noundef {{%.+}})
+// CHECK-NO-NANS: [[CALL_RES:%.+]] = call nnan nofpclass(nan) float @fn(float noundef nofpclass(nan) {{%.+}})
 // CHECK-NO-NANS: {{%.+}} = fadd nnan float {{%.+}}, [[CALL_RES]]
 
-// CHECK-NO-INFS: [[CALL_RES:%.+]] = call ninf float @fn(float noundef {{%.+}})
+// CHECK-NO-INFS: [[CALL_RES:%.+]] = call ninf nofpclass(inf) float @fn(float noundef nofpclass(inf) {{%.+}})
 // CHECK-NO-INFS: {{%.+}} = fadd ninf float {{%.+}}, [[CALL_RES]]
 
-// CHECK-FINITE: [[CALL_RES:%.+]] = call nnan ninf float @fn(float noundef {{%.+}})
+// CHECK-FINITE: [[CALL_RES:%.+]] = call nnan ninf nofpclass(nan inf) float @fn(float noundef nofpclass(nan inf) {{%.+}})
 // CHECK-FINITE: {{%.+}} = fadd nnan ninf float {{%.+}}, [[CALL_RES]]
 
 // CHECK-NO-SIGNED-ZEROS: [[CALL_RES:%.+]] = call nsz float @fn(float noundef {{%.+}})
 // CHECK-NO-SIGNED-ZEROS: {{%.+}} = fadd nsz float {{%.+}}, [[CALL_RES]]
 
 // CHECK-REASSOC: [[CALL_RES:%.+]] = call reassoc float @fn(float noundef {{%.+}})
 // CHECK-REASSOC: {{%.+}} = fadd reassoc float {{%.+}}, [[CALL_RES]]
 
 // CHECK-RECIP: [[CALL_RES:%.+]] = call arcp float @fn(float noundef {{%.+}})
 // CHECK-RECIP: {{%.+}} = fadd arcp float {{%.+}}, [[CALL_RES]]
 
 // CHECK-AFN: [[CALL_RES:%.+]] = call afn float @fn(float noundef {{%.+}})
 // CHECK-AFN: {{%.+}} = fadd afn float {{%.+}}, [[CALL_RES]]
 
 // CHECK-UNSAFE: [[CALL_RES:%.+]] = call reassoc nsz arcp afn float @fn(float noundef {{%.+}})
 // CHECK-UNSAFE: {{%.+}} = fadd reassoc nsz arcp afn float {{%.+}}, [[CALL_RES]]
 
-// CHECK-FAST: [[CALL_RES:%.+]] = call reassoc nnan ninf nsz arcp afn float @fn(float noundef {{%.+}})
+// CHECK-FAST: [[CALL_RES:%.+]] = call reassoc nnan ninf nsz arcp afn nofpclass(nan inf) float @fn(float noundef nofpclass(nan inf) {{%.+}})
 // CHECK-FAST: {{%.+}} = fadd reassoc nnan ninf nsz arcp afn float {{%.+}}, [[CALL_RES]]

clang/test/CodeGen/func-attr.c

 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffast-math \
 // RUN: -ffp-contract=fast -emit-llvm -o - %s | \
-// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE
+// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE,FINITEONLY
 
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
 // RUN: -ffp-contract=fast -emit-llvm -o - %s | \
-// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE
+// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-UNSAFE,NOFINITEONLY
 
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
 // RUN: -ffp-contract=on -emit-llvm -o - %s | \
-// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE
+// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE,NOFINITEONLY
 
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -funsafe-math-optimizations \
 // RUN: -ffp-contract=off -emit-llvm -o - %s | \
-// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE
+// RUN: FileCheck %s --check-prefixes=CHECK,CHECK-NOUNSAFE,NOFINITEONLY
 
 float foo(float a, float b) {
   return a+b;
 }
 
-// CHECK:              define{{.*}} float @foo(float noundef %{{.*}}, float noundef %{{.*}}){{.*}} [[ATTRS:#[0-9]+]]
+// FINITEONLY:    define{{.*}} float @foo(float noundef nofpclass(nan inf) %{{.*}}, float noundef nofpclass(nan inf) %{{.*}}){{.*}} [[ATTRS:#[0-9]+]]
+// NOFINITEONLY:    define{{.*}} float @foo(float noundef %{{.*}}, float noundef %{{.*}}){{.*}} [[ATTRS:#[0-9]+]]
+
 // CHECK:              attributes [[ATTRS]] = {
 // CHECK-SAME:           "approx-func-fp-math"="true"
 // CHECK-SAME:           "no-signed-zeros-fp-math"="true"
 // CHECK-SAME:           "no-trapping-math"="true"
 // CHECK-UNSAFE-SAME:    "unsafe-fp-math"="true"
 // CHECK-NOUNSAFE-NOT:   "unsafe-fp-math"="true"
 // CHECK-SAME:         }

clang/test/CodeGen/matrix-type-operators-fast-math.c

void subtract_compound_assign_matrix_double(dx5x5_t a, dx5x5_t b) {
   // 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]]
   // CHECK-NEXT:  store <25 x double> [[RES]], ptr {{.*}}, align 8
 
   a -= 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 %vf)
+  // CHECK-LABEL: define{{.*}} void @add_matrix_scalar_double_float(<25 x double> noundef %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
   // CHECK-NEXT:  [[SCALAR_EMBED:%.*]] = insertelement <25 x double> poison, double [[SCALAR_EXT]], i64 0
   // CHECK-NEXT:  [[SCALAR_EMBED1:%.*]] = shufflevector <25 x double> [[SCALAR_EMBED]], <25 x double> poison, <25 x i32> zeroinitializer
   // CHECK-NEXT:  [[RES:%.*]] = fadd reassoc nnan ninf nsz arcp afn <25 x double> [[MATRIX]], [[SCALAR_EMBED1]]
   // CHECK-NEXT:  store <25 x double> [[RES]], ptr {{.*}}, align 8
 
   a = a + 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 %vf)
+  // CHECK-LABEL: define{{.*}} void @add_compound_matrix_scalar_double_float(<25 x double> noundef %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
   // CHECK-NEXT:  [[SCALAR_EMBED:%.*]] = insertelement <25 x double> poison, double [[SCALAR_EXT]], i64 0
   // CHECK-NEXT:  [[SCALAR_EMBED1:%.*]] = shufflevector <25 x double> [[SCALAR_EMBED]], <25 x double> poison, <25 x i32> zeroinitializer
   // CHECK-NEXT:  [[RES:%.*]] = fadd reassoc nnan ninf nsz arcp afn <25 x double> [[MATRIX]], [[SCALAR_EMBED1]]
   // CHECK-NEXT:  store <25 x double> [[RES]], ptr {{.*}}, align 8
 
   a += 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 %vf)
+  // CHECK-LABEL: define{{.*}} void @subtract_compound_matrix_scalar_double_float(<25 x double> noundef %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
   // CHECK-NEXT:  [[SCALAR_EMBED:%.*]] = insertelement <25 x double> poison, double [[SCALAR_EXT]], i64 0
   // CHECK-NEXT:  [[SCALAR_EMBED1:%.*]] = shufflevector <25 x double> [[SCALAR_EMBED]], <25 x double> poison, <25 x i32> zeroinitializer
   // CHECK-NEXT:  [[RES:%.*]] = fsub reassoc nnan ninf nsz arcp afn <25 x double> [[MATRIX]], [[SCALAR_EMBED1]]
   // CHECK-NEXT:  store <25 x double> [[RES]], ptr {{.*}}, align 8
 

clang/test/CodeGen/nofpclass.c

 extern float extern_func(float, double, _Float16);
 extern float2 extern_func_vec(float2, double2, half2);
 extern _Complex float extern_complex(_Complex float, _Complex double, _Complex _Float16);
 extern float variadic(float, ...);
 extern dx5x5_t extern_matrix(dx5x5_t);
 
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local float @defined_func_f32
-// CFINITEONLY-SAME: (float noundef [[A:%.*]], float noundef [[B:%.*]], float noundef [[C:%.*]]) #[[ATTR0:[0-9]+]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @defined_func_f32
+// CFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[A:%.*]], float noundef nofpclass(nan inf) [[B:%.*]], float noundef nofpclass(nan inf) [[C:%.*]]) #[[ATTR0:[0-9]+]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    [[B_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    [[C_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // CFINITEONLY-NEXT:    store float [[B]], ptr [[B_ADDR]], align 4
 // CFINITEONLY-NEXT:    store float [[C]], ptr [[C_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = load float, ptr [[B_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP2:%.*]] = load float, ptr [[C_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP3:%.*]] = call nnan ninf float @llvm.fma.f32(float [[TMP0]], float [[TMP1]], float [[TMP2]])
 // CFINITEONLY-NEXT:    [[ADD:%.*]] = fadd nnan ninf float [[TMP3]], 4.000000e+00
 // CFINITEONLY-NEXT:    ret float [[ADD]]
 //
 // CLFINITEONLY: Function Attrs: mustprogress nofree norecurse nosync nounwind willreturn memory(none)
-// CLFINITEONLY-LABEL: define dso_local float @defined_func_f32
-// CLFINITEONLY-SAME: (float noundef [[A:%.*]], float noundef [[B:%.*]], float noundef [[C:%.*]]) local_unnamed_addr #[[ATTR0:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @defined_func_f32
+// CLFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[A:%.*]], float noundef nofpclass(nan inf) [[B:%.*]], float noundef nofpclass(nan inf) [[C:%.*]]) local_unnamed_addr #[[ATTR0:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[TMP0:%.*]] = tail call nnan ninf float @llvm.fma.f32(float [[A]], float [[B]], float [[C]])
 // CLFINITEONLY-NEXT:    [[ADD:%.*]] = fadd nnan ninf float [[TMP0]], 4.000000e+00
 // CLFINITEONLY-NEXT:    ret float [[ADD]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local float @defined_func_f32
-// NONANS-SAME: (float noundef [[A:%.*]], float noundef [[B:%.*]], float noundef [[C:%.*]]) #[[ATTR0:[0-9]+]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) float @defined_func_f32
+// NONANS-SAME: (float noundef nofpclass(nan) [[A:%.*]], float noundef nofpclass(nan) [[B:%.*]], float noundef nofpclass(nan) [[C:%.*]]) #[[ATTR0:[0-9]+]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    [[B_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    [[C_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // NONANS-NEXT:    store float [[B]], ptr [[B_ADDR]], align 4
 // NONANS-NEXT:    store float [[C]], ptr [[C_ADDR]], align 4
 // NONANS-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // NONANS-NEXT:    [[TMP1:%.*]] = load float, ptr [[B_ADDR]], align 4
 // NONANS-NEXT:    [[TMP2:%.*]] = load float, ptr [[C_ADDR]], align 4
 // NONANS-NEXT:    [[TMP3:%.*]] = call nnan float @llvm.fma.f32(float [[TMP0]], float [[TMP1]], float [[TMP2]])
 // NONANS-NEXT:    [[ADD:%.*]] = fadd nnan float [[TMP3]], 4.000000e+00
 // NONANS-NEXT:    ret float [[ADD]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local float @defined_func_f32
-// NOINFS-SAME: (float noundef [[A:%.*]], float noundef [[B:%.*]], float noundef [[C:%.*]]) #[[ATTR0:[0-9]+]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) float @defined_func_f32
+// NOINFS-SAME: (float noundef nofpclass(inf) [[A:%.*]], float noundef nofpclass(inf) [[B:%.*]], float noundef nofpclass(inf) [[C:%.*]]) #[[ATTR0:[0-9]+]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    [[B_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    [[C_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // NOINFS-NEXT:    store float [[B]], ptr [[B_ADDR]], align 4
 // NOINFS-NEXT:    store float [[C]], ptr [[C_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP1:%.*]] = load float, ptr [[B_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP2:%.*]] = load float, ptr [[C_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP3:%.*]] = call ninf float @llvm.fma.f32(float [[TMP0]], float [[TMP1]], float [[TMP2]])
 // NOINFS-NEXT:    [[ADD:%.*]] = fadd ninf float [[TMP3]], 4.000000e+00
 // NOINFS-NEXT:    ret float [[ADD]]
 //
 float defined_func_f32(float a, float b, float c) {
   return __builtin_fmaf(a, b, c) + 4.0f;
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local <2 x double> @defined_func_v2f64
-// CFINITEONLY-SAME: (<2 x double> noundef [[A:%.*]], <2 x double> noundef [[B:%.*]], <2 x double> noundef [[C:%.*]]) #[[ATTR2:[0-9]+]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x double> @defined_func_v2f64
+// CFINITEONLY-SAME: (<2 x double> noundef nofpclass(nan inf) [[A:%.*]], <2 x double> noundef nofpclass(nan inf) [[B:%.*]], <2 x double> noundef nofpclass(nan inf) [[C:%.*]]) #[[ATTR2:[0-9]+]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[A_ADDR:%.*]] = alloca <2 x double>, align 16
 // CFINITEONLY-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // CFINITEONLY-NEXT:    [[C_ADDR:%.*]] = alloca <2 x double>, align 16
 // CFINITEONLY-NEXT:    store <2 x double> [[A]], ptr [[A_ADDR]], align 16
 // CFINITEONLY-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // CFINITEONLY-NEXT:    store <2 x double> [[C]], ptr [[C_ADDR]], align 16
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load <2 x double>, ptr [[A_ADDR]], align 16
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // CFINITEONLY-NEXT:    [[TMP2:%.*]] = load <2 x double>, ptr [[C_ADDR]], align 16
 // CFINITEONLY-NEXT:    [[TMP3:%.*]] = call nnan ninf <2 x double> @llvm.fma.v2f64(<2 x double> [[TMP0]], <2 x double> [[TMP1]], <2 x double> [[TMP2]])
 // CFINITEONLY-NEXT:    [[ADD:%.*]] = fadd nnan ninf <2 x double> [[TMP3]], <double 4.000000e+00, double 4.000000e+00>
 // CFINITEONLY-NEXT:    ret <2 x double> [[ADD]]
 //
 // CLFINITEONLY: Function Attrs: mustprogress nofree norecurse nosync nounwind willreturn memory(none)
-// CLFINITEONLY-LABEL: define dso_local <2 x double> @defined_func_v2f64
-// CLFINITEONLY-SAME: (<2 x double> noundef [[A:%.*]], <2 x double> noundef [[B:%.*]], <2 x double> noundef [[C:%.*]]) local_unnamed_addr #[[ATTR2:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x double> @defined_func_v2f64
+// CLFINITEONLY-SAME: (<2 x double> noundef nofpclass(nan inf) [[A:%.*]], <2 x double> noundef nofpclass(nan inf) [[B:%.*]], <2 x double> noundef nofpclass(nan inf) [[C:%.*]]) local_unnamed_addr #[[ATTR2:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[TMP0:%.*]] = tail call nnan ninf <2 x double> @llvm.fma.v2f64(<2 x double> [[A]], <2 x double> [[B]], <2 x double> [[C]])
 // CLFINITEONLY-NEXT:    [[ADD:%.*]] = fadd nnan ninf <2 x double> [[TMP0]], <double 4.000000e+00, double 4.000000e+00>
 // CLFINITEONLY-NEXT:    ret <2 x double> [[ADD]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local <2 x double> @defined_func_v2f64
-// NONANS-SAME: (<2 x double> noundef [[A:%.*]], <2 x double> noundef [[B:%.*]], <2 x double> noundef [[C:%.*]]) #[[ATTR2:[0-9]+]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) <2 x double> @defined_func_v2f64
+// NONANS-SAME: (<2 x double> noundef nofpclass(nan) [[A:%.*]], <2 x double> noundef nofpclass(nan) [[B:%.*]], <2 x double> noundef nofpclass(nan) [[C:%.*]]) #[[ATTR2:[0-9]+]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[A_ADDR:%.*]] = alloca <2 x double>, align 16
 // NONANS-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // NONANS-NEXT:    [[C_ADDR:%.*]] = alloca <2 x double>, align 16
 // NONANS-NEXT:    store <2 x double> [[A]], ptr [[A_ADDR]], align 16
 // NONANS-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // NONANS-NEXT:    store <2 x double> [[C]], ptr [[C_ADDR]], align 16
 // NONANS-NEXT:    [[TMP0:%.*]] = load <2 x double>, ptr [[A_ADDR]], align 16
 // NONANS-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // NONANS-NEXT:    [[TMP2:%.*]] = load <2 x double>, ptr [[C_ADDR]], align 16
 // NONANS-NEXT:    [[TMP3:%.*]] = call nnan <2 x double> @llvm.fma.v2f64(<2 x double> [[TMP0]], <2 x double> [[TMP1]], <2 x double> [[TMP2]])
 // NONANS-NEXT:    [[ADD:%.*]] = fadd nnan <2 x double> [[TMP3]], <double 4.000000e+00, double 4.000000e+00>
 // NONANS-NEXT:    ret <2 x double> [[ADD]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local <2 x double> @defined_func_v2f64
-// NOINFS-SAME: (<2 x double> noundef [[A:%.*]], <2 x double> noundef [[B:%.*]], <2 x double> noundef [[C:%.*]]) #[[ATTR2:[0-9]+]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) <2 x double> @defined_func_v2f64
+// NOINFS-SAME: (<2 x double> noundef nofpclass(inf) [[A:%.*]], <2 x double> noundef nofpclass(inf) [[B:%.*]], <2 x double> noundef nofpclass(inf) [[C:%.*]]) #[[ATTR2:[0-9]+]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[A_ADDR:%.*]] = alloca <2 x double>, align 16
 // NOINFS-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // NOINFS-NEXT:    [[C_ADDR:%.*]] = alloca <2 x double>, align 16
 // NOINFS-NEXT:    store <2 x double> [[A]], ptr [[A_ADDR]], align 16
 // NOINFS-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // NOINFS-NEXT:    store <2 x double> [[C]], ptr [[C_ADDR]], align 16
 // NOINFS-NEXT:    [[TMP0:%.*]] = load <2 x double>, ptr [[A_ADDR]], align 16
 // NOINFS-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // NOINFS-NEXT:    [[TMP2:%.*]] = load <2 x double>, ptr [[C_ADDR]], align 16
 // NOINFS-NEXT:    [[TMP3:%.*]] = call ninf <2 x double> @llvm.fma.v2f64(<2 x double> [[TMP0]], <2 x double> [[TMP1]], <2 x double> [[TMP2]])
 // NOINFS-NEXT:    [[ADD:%.*]] = fadd ninf <2 x double> [[TMP3]], <double 4.000000e+00, double 4.000000e+00>
 // NOINFS-NEXT:    ret <2 x double> [[ADD]]
 //
 double2 defined_func_v2f64(double2 a, double2 b, double2 c) {
   return __builtin_elementwise_fma(a, b, c) + 4.0;
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local float @call_extern_func
-// CFINITEONLY-SAME: (float noundef [[A:%.*]], double noundef [[B:%.*]], half noundef [[C:%.*]]) #[[ATTR0]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_extern_func
+// CFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[A:%.*]], double noundef nofpclass(nan inf) [[B:%.*]], half noundef nofpclass(nan inf) [[C:%.*]]) #[[ATTR0]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    [[B_ADDR:%.*]] = alloca double, align 8
 // CFINITEONLY-NEXT:    [[C_ADDR:%.*]] = alloca half, align 2
 // CFINITEONLY-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // CFINITEONLY-NEXT:    store double [[B]], ptr [[B_ADDR]], align 8
 // CFINITEONLY-NEXT:    store half [[C]], ptr [[C_ADDR]], align 2
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = load double, ptr [[B_ADDR]], align 8
 // CFINITEONLY-NEXT:    [[TMP2:%.*]] = load half, ptr [[C_ADDR]], align 2
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf float @extern_func(float noundef [[TMP0]], double noundef [[TMP1]], half noundef [[TMP2]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) float @extern_func(float noundef nofpclass(nan inf) [[TMP0]], double noundef nofpclass(nan inf) [[TMP1]], half noundef nofpclass(nan inf) [[TMP2]])
 // CFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local float @call_extern_func
-// CLFINITEONLY-SAME: (float noundef [[A:%.*]], double noundef [[B:%.*]], half noundef [[C:%.*]]) local_unnamed_addr #[[ATTR3:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_extern_func
+// CLFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[A:%.*]], double noundef nofpclass(nan inf) [[B:%.*]], half noundef nofpclass(nan inf) [[C:%.*]]) local_unnamed_addr #[[ATTR3:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf float @extern_func(float noundef [[A]], double noundef [[B]], half noundef [[C]]) #[[ATTR10:[0-9]+]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) float @extern_func(float noundef nofpclass(nan inf) [[A]], double noundef nofpclass(nan inf) [[B]], half noundef nofpclass(nan inf) [[C]]) #[[ATTR10:[0-9]+]]
 // CLFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local float @call_extern_func
-// NONANS-SAME: (float noundef [[A:%.*]], double noundef [[B:%.*]], half noundef [[C:%.*]]) #[[ATTR0]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) float @call_extern_func
+// NONANS-SAME: (float noundef nofpclass(nan) [[A:%.*]], double noundef nofpclass(nan) [[B:%.*]], half noundef nofpclass(nan) [[C:%.*]]) #[[ATTR0]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    [[B_ADDR:%.*]] = alloca double, align 8
 // NONANS-NEXT:    [[C_ADDR:%.*]] = alloca half, align 2
 // NONANS-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // NONANS-NEXT:    store double [[B]], ptr [[B_ADDR]], align 8
 // NONANS-NEXT:    store half [[C]], ptr [[C_ADDR]], align 2
 // NONANS-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // NONANS-NEXT:    [[TMP1:%.*]] = load double, ptr [[B_ADDR]], align 8
 // NONANS-NEXT:    [[TMP2:%.*]] = load half, ptr [[C_ADDR]], align 2
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan float @extern_func(float noundef [[TMP0]], double noundef [[TMP1]], half noundef [[TMP2]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) float @extern_func(float noundef nofpclass(nan) [[TMP0]], double noundef nofpclass(nan) [[TMP1]], half noundef nofpclass(nan) [[TMP2]])
 // NONANS-NEXT:    ret float [[CALL]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local float @call_extern_func
-// NOINFS-SAME: (float noundef [[A:%.*]], double noundef [[B:%.*]], half noundef [[C:%.*]]) #[[ATTR0]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) float @call_extern_func
+// NOINFS-SAME: (float noundef nofpclass(inf) [[A:%.*]], double noundef nofpclass(inf) [[B:%.*]], half noundef nofpclass(inf) [[C:%.*]]) #[[ATTR0]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    [[B_ADDR:%.*]] = alloca double, align 8
 // NOINFS-NEXT:    [[C_ADDR:%.*]] = alloca half, align 2
 // NOINFS-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 // NOINFS-NEXT:    store double [[B]], ptr [[B_ADDR]], align 8
 // NOINFS-NEXT:    store half [[C]], ptr [[C_ADDR]], align 2
 // NOINFS-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP1:%.*]] = load double, ptr [[B_ADDR]], align 8
 // NOINFS-NEXT:    [[TMP2:%.*]] = load half, ptr [[C_ADDR]], align 2
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf float @extern_func(float noundef [[TMP0]], double noundef [[TMP1]], half noundef [[TMP2]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) float @extern_func(float noundef nofpclass(inf) [[TMP0]], double noundef nofpclass(inf) [[TMP1]], half noundef nofpclass(inf) [[TMP2]])
 // NOINFS-NEXT:    ret float [[CALL]]
 //
 float call_extern_func(float a, double b, _Float16 c) {
   return extern_func(a, b, c);
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local double @call_extern_func_vec
-// CFINITEONLY-SAME: (double noundef [[A_COERCE:%.*]], <2 x double> noundef [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) double @call_extern_func_vec
+// CFINITEONLY-SAME: (double noundef nofpclass(nan inf) [[A_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[RETVAL:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[A:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[C:%.*]] = alloca <2 x half>, align 4
 // CFINITEONLY-NEXT:    [[A_ADDR:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // CFINITEONLY-NEXT:    [[C_ADDR:%.*]] = alloca <2 x half>, align 4
 // CFINITEONLY-NEXT:    [[COERCE:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[COERCE3:%.*]] = alloca <2 x half>, align 4
 // CFINITEONLY-NEXT:    [[COERCE4:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    store double [[A_COERCE]], ptr [[A]], align 8
 // CFINITEONLY-NEXT:    [[A1:%.*]] = load <2 x float>, ptr [[A]], align 8
 // CFINITEONLY-NEXT:    store i32 [[C_COERCE]], ptr [[C]], align 4
 // CFINITEONLY-NEXT:    [[C2:%.*]] = load <2 x half>, ptr [[C]], align 4
 // CFINITEONLY-NEXT:    store <2 x float> [[A1]], ptr [[A_ADDR]], align 8
 // CFINITEONLY-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // CFINITEONLY-NEXT:    store <2 x half> [[C2]], ptr [[C_ADDR]], align 4
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load <2 x float>, ptr [[A_ADDR]], align 8
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // CFINITEONLY-NEXT:    [[TMP2:%.*]] = load <2 x half>, ptr [[C_ADDR]], align 4
 // CFINITEONLY-NEXT:    store <2 x float> [[TMP0]], ptr [[COERCE]], align 8
 // CFINITEONLY-NEXT:    [[TMP3:%.*]] = load double, ptr [[COERCE]], align 8
 // CFINITEONLY-NEXT:    store <2 x half> [[TMP2]], ptr [[COERCE3]], align 4
 // CFINITEONLY-NEXT:    [[TMP4:%.*]] = load i32, ptr [[COERCE3]], align 4
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf double @extern_func_vec(double noundef [[TMP3]], <2 x double> noundef [[TMP1]], i32 noundef [[TMP4]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) double @extern_func_vec(double noundef nofpclass(nan inf) [[TMP3]], <2 x double> noundef nofpclass(nan inf) [[TMP1]], i32 noundef [[TMP4]])
 // CFINITEONLY-NEXT:    store double [[CALL]], ptr [[COERCE4]], align 8
 // CFINITEONLY-NEXT:    [[TMP5:%.*]] = load <2 x float>, ptr [[COERCE4]], align 8
 // CFINITEONLY-NEXT:    store <2 x float> [[TMP5]], ptr [[RETVAL]], align 8
 // CFINITEONLY-NEXT:    [[TMP6:%.*]] = load double, ptr [[RETVAL]], align 8
 // CFINITEONLY-NEXT:    ret double [[TMP6]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local double @call_extern_func_vec
-// CLFINITEONLY-SAME: (double noundef [[A_COERCE:%.*]], <2 x double> noundef [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR5:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) double @call_extern_func_vec
+// CLFINITEONLY-SAME: (double noundef nofpclass(nan inf) [[A_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR5:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf double @extern_func_vec(double noundef [[A_COERCE]], <2 x double> noundef [[B]], i32 noundef [[C_COERCE]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) double @extern_func_vec(double noundef nofpclass(nan inf) [[A_COERCE]], <2 x double> noundef nofpclass(nan inf) [[B]], i32 noundef [[C_COERCE]]) #[[ATTR10]]
 // CLFINITEONLY-NEXT:    ret double [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local double @call_extern_func_vec
-// NONANS-SAME: (double noundef [[A_COERCE:%.*]], <2 x double> noundef [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) double @call_extern_func_vec
+// NONANS-SAME: (double noundef nofpclass(nan) [[A_COERCE:%.*]], <2 x double> noundef nofpclass(nan) [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[RETVAL:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[A:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[C:%.*]] = alloca <2 x half>, align 4
 // NONANS-NEXT:    [[A_ADDR:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // NONANS-NEXT:    [[C_ADDR:%.*]] = alloca <2 x half>, align 4
 // NONANS-NEXT:    [[COERCE:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[COERCE3:%.*]] = alloca <2 x half>, align 4
 // NONANS-NEXT:    [[COERCE4:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    store double [[A_COERCE]], ptr [[A]], align 8
 // NONANS-NEXT:    [[A1:%.*]] = load <2 x float>, ptr [[A]], align 8
 // NONANS-NEXT:    store i32 [[C_COERCE]], ptr [[C]], align 4
 // NONANS-NEXT:    [[C2:%.*]] = load <2 x half>, ptr [[C]], align 4
 // NONANS-NEXT:    store <2 x float> [[A1]], ptr [[A_ADDR]], align 8
 // NONANS-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // NONANS-NEXT:    store <2 x half> [[C2]], ptr [[C_ADDR]], align 4
 // NONANS-NEXT:    [[TMP0:%.*]] = load <2 x float>, ptr [[A_ADDR]], align 8
 // NONANS-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // NONANS-NEXT:    [[TMP2:%.*]] = load <2 x half>, ptr [[C_ADDR]], align 4
 // NONANS-NEXT:    store <2 x float> [[TMP0]], ptr [[COERCE]], align 8
 // NONANS-NEXT:    [[TMP3:%.*]] = load double, ptr [[COERCE]], align 8
 // NONANS-NEXT:    store <2 x half> [[TMP2]], ptr [[COERCE3]], align 4
 // NONANS-NEXT:    [[TMP4:%.*]] = load i32, ptr [[COERCE3]], align 4
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan double @extern_func_vec(double noundef [[TMP3]], <2 x double> noundef [[TMP1]], i32 noundef [[TMP4]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) double @extern_func_vec(double noundef nofpclass(nan) [[TMP3]], <2 x double> noundef nofpclass(nan) [[TMP1]], i32 noundef [[TMP4]])
 // NONANS-NEXT:    store double [[CALL]], ptr [[COERCE4]], align 8
 // NONANS-NEXT:    [[TMP5:%.*]] = load <2 x float>, ptr [[COERCE4]], align 8
 // NONANS-NEXT:    store <2 x float> [[TMP5]], ptr [[RETVAL]], align 8
 // NONANS-NEXT:    [[TMP6:%.*]] = load double, ptr [[RETVAL]], align 8
 // NONANS-NEXT:    ret double [[TMP6]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local double @call_extern_func_vec
-// NOINFS-SAME: (double noundef [[A_COERCE:%.*]], <2 x double> noundef [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) double @call_extern_func_vec
+// NOINFS-SAME: (double noundef nofpclass(inf) [[A_COERCE:%.*]], <2 x double> noundef nofpclass(inf) [[B:%.*]], i32 noundef [[C_COERCE:%.*]]) #[[ATTR2]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[RETVAL:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[A:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[C:%.*]] = alloca <2 x half>, align 4
 // NOINFS-NEXT:    [[A_ADDR:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[B_ADDR:%.*]] = alloca <2 x double>, align 16
 // NOINFS-NEXT:    [[C_ADDR:%.*]] = alloca <2 x half>, align 4
 // NOINFS-NEXT:    [[COERCE:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[COERCE3:%.*]] = alloca <2 x half>, align 4
 // NOINFS-NEXT:    [[COERCE4:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    store double [[A_COERCE]], ptr [[A]], align 8
 // NOINFS-NEXT:    [[A1:%.*]] = load <2 x float>, ptr [[A]], align 8
 // NOINFS-NEXT:    store i32 [[C_COERCE]], ptr [[C]], align 4
 // NOINFS-NEXT:    [[C2:%.*]] = load <2 x half>, ptr [[C]], align 4
 // NOINFS-NEXT:    store <2 x float> [[A1]], ptr [[A_ADDR]], align 8
 // NOINFS-NEXT:    store <2 x double> [[B]], ptr [[B_ADDR]], align 16
 // NOINFS-NEXT:    store <2 x half> [[C2]], ptr [[C_ADDR]], align 4
 // NOINFS-NEXT:    [[TMP0:%.*]] = load <2 x float>, ptr [[A_ADDR]], align 8
 // NOINFS-NEXT:    [[TMP1:%.*]] = load <2 x double>, ptr [[B_ADDR]], align 16
 // NOINFS-NEXT:    [[TMP2:%.*]] = load <2 x half>, ptr [[C_ADDR]], align 4
 // NOINFS-NEXT:    store <2 x float> [[TMP0]], ptr [[COERCE]], align 8
 // NOINFS-NEXT:    [[TMP3:%.*]] = load double, ptr [[COERCE]], align 8
 // NOINFS-NEXT:    store <2 x half> [[TMP2]], ptr [[COERCE3]], align 4
 // NOINFS-NEXT:    [[TMP4:%.*]] = load i32, ptr [[COERCE3]], align 4
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf double @extern_func_vec(double noundef [[TMP3]], <2 x double> noundef [[TMP1]], i32 noundef [[TMP4]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) double @extern_func_vec(double noundef nofpclass(inf) [[TMP3]], <2 x double> noundef nofpclass(inf) [[TMP1]], i32 noundef [[TMP4]])
 // NOINFS-NEXT:    store double [[CALL]], ptr [[COERCE4]], align 8
 // NOINFS-NEXT:    [[TMP5:%.*]] = load <2 x float>, ptr [[COERCE4]], align 8
 // NOINFS-NEXT:    store <2 x float> [[TMP5]], ptr [[RETVAL]], align 8
 // NOINFS-NEXT:    [[TMP6:%.*]] = load double, ptr [[RETVAL]], align 8
 // NOINFS-NEXT:    ret double [[TMP6]]
 //
 float2 call_extern_func_vec(float2 a, double2 b, half2 c) {
   return extern_func_vec(a, b, c);
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local <2 x float> @defined_complex_func
-// CFINITEONLY-SAME: (<2 x float> noundef [[A_COERCE:%.*]], double noundef [[B_COERCE0:%.*]], double noundef [[B_COERCE1:%.*]], <2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x float> @defined_complex_func
+// CFINITEONLY-SAME: (<2 x float> noundef nofpclass(nan inf) [[A_COERCE:%.*]], double noundef nofpclass(nan inf) [[B_COERCE0:%.*]], double noundef nofpclass(nan inf) [[B_COERCE1:%.*]], <2 x half> noundef nofpclass(nan inf) [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[RETVAL:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    [[A:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    [[B:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // CFINITEONLY-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    [[COERCE1:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[COERCE2:%.*]] = alloca { half, half }, align 2

 // CFINITEONLY-NEXT:    [[TMP4:%.*]] = load double, ptr [[TMP3]], align 8
 // CFINITEONLY-NEXT:    [[TMP5:%.*]] = getelementptr inbounds { double, double }, ptr [[COERCE1]], i32 0, i32 1
 // CFINITEONLY-NEXT:    [[TMP6:%.*]] = load double, ptr [[TMP5]], align 8
 // CFINITEONLY-NEXT:    [[COERCE2_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[COERCE2_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store half [[C_REAL]], ptr [[COERCE2_REALP]], align 2
 // CFINITEONLY-NEXT:    store half [[C_IMAG]], ptr [[COERCE2_IMAGP]], align 2
 // CFINITEONLY-NEXT:    [[TMP7:%.*]] = load <2 x half>, ptr [[COERCE2]], align 2
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf <2 x float> @extern_complex(<2 x float> noundef [[TMP2]], double noundef [[TMP4]], double noundef [[TMP6]], <2 x half> noundef [[TMP7]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) <2 x float> @extern_complex(<2 x float> noundef nofpclass(nan inf) [[TMP2]], double noundef nofpclass(nan inf) [[TMP4]], double noundef nofpclass(nan inf) [[TMP6]], <2 x half> noundef nofpclass(nan inf) [[TMP7]])
 // CFINITEONLY-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE3]], align 4
 // CFINITEONLY-NEXT:    [[COERCE3_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[COERCE3_REAL:%.*]] = load float, ptr [[COERCE3_REALP]], align 4
 // CFINITEONLY-NEXT:    [[COERCE3_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 1
 // CFINITEONLY-NEXT:    [[COERCE3_IMAG:%.*]] = load float, ptr [[COERCE3_IMAGP]], align 4
 // CFINITEONLY-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store float [[COERCE3_REAL]], ptr [[RETVAL_REALP]], align 4
 // CFINITEONLY-NEXT:    store float [[COERCE3_IMAG]], ptr [[RETVAL_IMAGP]], align 4
 // CFINITEONLY-NEXT:    [[TMP8:%.*]] = load <2 x float>, ptr [[RETVAL]], align 4
 // CFINITEONLY-NEXT:    ret <2 x float> [[TMP8]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local <2 x float> @defined_complex_func
-// CLFINITEONLY-SAME: (<2 x float> noundef [[A_COERCE:%.*]], double noundef [[B_COERCE0:%.*]], double noundef [[B_COERCE1:%.*]], <2 x half> noundef [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR6:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x float> @defined_complex_func
+// CLFINITEONLY-SAME: (<2 x float> noundef nofpclass(nan inf) [[A_COERCE:%.*]], double noundef nofpclass(nan inf) [[B_COERCE0:%.*]], double noundef nofpclass(nan inf) [[B_COERCE1:%.*]], <2 x half> noundef nofpclass(nan inf) [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR6:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf <2 x float> @extern_complex(<2 x float> noundef [[A_COERCE]], double noundef [[B_COERCE0]], double noundef [[B_COERCE1]], <2 x half> noundef [[C_COERCE]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) <2 x float> @extern_complex(<2 x float> noundef nofpclass(nan inf) [[A_COERCE]], double noundef nofpclass(nan inf) [[B_COERCE0]], double noundef nofpclass(nan inf) [[B_COERCE1]], <2 x half> noundef nofpclass(nan inf) [[C_COERCE]]) #[[ATTR10]]
 // CLFINITEONLY-NEXT:    ret <2 x float> [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local <2 x float> @defined_complex_func
-// NONANS-SAME: (<2 x float> noundef [[A_COERCE:%.*]], double noundef [[B_COERCE0:%.*]], double noundef [[B_COERCE1:%.*]], <2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) <2 x float> @defined_complex_func
+// NONANS-SAME: (<2 x float> noundef nofpclass(nan) [[A_COERCE:%.*]], double noundef nofpclass(nan) [[B_COERCE0:%.*]], double noundef nofpclass(nan) [[B_COERCE1:%.*]], <2 x half> noundef nofpclass(nan) [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[RETVAL:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    [[A:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    [[B:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // NONANS-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    [[COERCE1:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[COERCE2:%.*]] = alloca { half, half }, align 2

 // NONANS-NEXT:    [[TMP4:%.*]] = load double, ptr [[TMP3]], align 8
 // NONANS-NEXT:    [[TMP5:%.*]] = getelementptr inbounds { double, double }, ptr [[COERCE1]], i32 0, i32 1
 // NONANS-NEXT:    [[TMP6:%.*]] = load double, ptr [[TMP5]], align 8
 // NONANS-NEXT:    [[COERCE2_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 0
 // NONANS-NEXT:    [[COERCE2_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 1
 // NONANS-NEXT:    store half [[C_REAL]], ptr [[COERCE2_REALP]], align 2
 // NONANS-NEXT:    store half [[C_IMAG]], ptr [[COERCE2_IMAGP]], align 2
 // NONANS-NEXT:    [[TMP7:%.*]] = load <2 x half>, ptr [[COERCE2]], align 2
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan <2 x float> @extern_complex(<2 x float> noundef [[TMP2]], double noundef [[TMP4]], double noundef [[TMP6]], <2 x half> noundef [[TMP7]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) <2 x float> @extern_complex(<2 x float> noundef nofpclass(nan) [[TMP2]], double noundef nofpclass(nan) [[TMP4]], double noundef nofpclass(nan) [[TMP6]], <2 x half> noundef nofpclass(nan) [[TMP7]])
 // NONANS-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE3]], align 4
 // NONANS-NEXT:    [[COERCE3_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 0
 // NONANS-NEXT:    [[COERCE3_REAL:%.*]] = load float, ptr [[COERCE3_REALP]], align 4
 // NONANS-NEXT:    [[COERCE3_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 1
 // NONANS-NEXT:    [[COERCE3_IMAG:%.*]] = load float, ptr [[COERCE3_IMAGP]], align 4
 // NONANS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 0
 // NONANS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 1
 // NONANS-NEXT:    store float [[COERCE3_REAL]], ptr [[RETVAL_REALP]], align 4
 // NONANS-NEXT:    store float [[COERCE3_IMAG]], ptr [[RETVAL_IMAGP]], align 4
 // NONANS-NEXT:    [[TMP8:%.*]] = load <2 x float>, ptr [[RETVAL]], align 4
 // NONANS-NEXT:    ret <2 x float> [[TMP8]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local <2 x float> @defined_complex_func
-// NOINFS-SAME: (<2 x float> noundef [[A_COERCE:%.*]], double noundef [[B_COERCE0:%.*]], double noundef [[B_COERCE1:%.*]], <2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) <2 x float> @defined_complex_func
+// NOINFS-SAME: (<2 x float> noundef nofpclass(inf) [[A_COERCE:%.*]], double noundef nofpclass(inf) [[B_COERCE0:%.*]], double noundef nofpclass(inf) [[B_COERCE1:%.*]], <2 x half> noundef nofpclass(inf) [[C_COERCE:%.*]]) #[[ATTR4:[0-9]+]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[RETVAL:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    [[A:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    [[B:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // NOINFS-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    [[COERCE1:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[COERCE2:%.*]] = alloca { half, half }, align 2

 // NOINFS-NEXT:    [[TMP4:%.*]] = load double, ptr [[TMP3]], align 8
 // NOINFS-NEXT:    [[TMP5:%.*]] = getelementptr inbounds { double, double }, ptr [[COERCE1]], i32 0, i32 1
 // NOINFS-NEXT:    [[TMP6:%.*]] = load double, ptr [[TMP5]], align 8
 // NOINFS-NEXT:    [[COERCE2_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 0
 // NOINFS-NEXT:    [[COERCE2_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE2]], i32 0, i32 1
 // NOINFS-NEXT:    store half [[C_REAL]], ptr [[COERCE2_REALP]], align 2
 // NOINFS-NEXT:    store half [[C_IMAG]], ptr [[COERCE2_IMAGP]], align 2
 // NOINFS-NEXT:    [[TMP7:%.*]] = load <2 x half>, ptr [[COERCE2]], align 2
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf <2 x float> @extern_complex(<2 x float> noundef [[TMP2]], double noundef [[TMP4]], double noundef [[TMP6]], <2 x half> noundef [[TMP7]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) <2 x float> @extern_complex(<2 x float> noundef nofpclass(inf) [[TMP2]], double noundef nofpclass(inf) [[TMP4]], double noundef nofpclass(inf) [[TMP6]], <2 x half> noundef nofpclass(inf) [[TMP7]])
 // NOINFS-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE3]], align 4
 // NOINFS-NEXT:    [[COERCE3_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 0
 // NOINFS-NEXT:    [[COERCE3_REAL:%.*]] = load float, ptr [[COERCE3_REALP]], align 4
 // NOINFS-NEXT:    [[COERCE3_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE3]], i32 0, i32 1
 // NOINFS-NEXT:    [[COERCE3_IMAG:%.*]] = load float, ptr [[COERCE3_IMAGP]], align 4
 // NOINFS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 0
 // NOINFS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[RETVAL]], i32 0, i32 1
 // NOINFS-NEXT:    store float [[COERCE3_REAL]], ptr [[RETVAL_REALP]], align 4
 // NOINFS-NEXT:    store float [[COERCE3_IMAG]], ptr [[RETVAL_IMAGP]], align 4
 // NOINFS-NEXT:    [[TMP8:%.*]] = load <2 x float>, ptr [[RETVAL]], align 4
 // NOINFS-NEXT:    ret <2 x float> [[TMP8]]
 //
 _Complex float defined_complex_func(_Complex float a, _Complex double b, _Complex _Float16 c) {
   return extern_complex(a, b, c);
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
 // CFINITEONLY-LABEL: define dso_local { double, double } @defined_complex_func_f64_ret
-// CFINITEONLY-SAME: (double noundef [[C_COERCE0:%.*]], double noundef [[C_COERCE1:%.*]]) #[[ATTR0]] {
+// CFINITEONLY-SAME: (double noundef nofpclass(nan inf) [[C_COERCE0:%.*]], double noundef nofpclass(nan inf) [[C_COERCE1:%.*]]) #[[ATTR0]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[RETVAL:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[C:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0
 // CFINITEONLY-NEXT:    store double [[C_COERCE0]], ptr [[TMP0]], align 8
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store double [[C_COERCE1]], ptr [[TMP1]], align 8
 // CFINITEONLY-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0

 // CFINITEONLY-NEXT:    [[MUL_R:%.*]] = fsub nnan ninf double [[MUL_AC]], [[MUL_BD]]
 // CFINITEONLY-NEXT:    [[MUL_I:%.*]] = fadd nnan ninf double [[MUL_AD]], [[MUL_BC]]
 // CFINITEONLY-NEXT:    [[ISNAN_CMP:%.*]] = fcmp nnan ninf uno double [[MUL_R]], [[MUL_R]]
 // CFINITEONLY-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2:![0-9]+]]
 // CFINITEONLY:       complex_mul_imag_nan:
 // CFINITEONLY-NEXT:    [[ISNAN_CMP5:%.*]] = fcmp nnan ninf uno double [[MUL_I]], [[MUL_I]]
 // CFINITEONLY-NEXT:    br i1 [[ISNAN_CMP5]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // CFINITEONLY:       complex_mul_libcall:
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef [[C_REAL]], double noundef [[C_IMAG]], double noundef [[C_REAL2]], double noundef [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef nofpclass(nan inf) [[C_REAL]], double noundef nofpclass(nan inf) [[C_IMAG]], double noundef nofpclass(nan inf) [[C_REAL2]], double noundef nofpclass(nan inf) [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
 // CFINITEONLY-NEXT:    [[TMP2:%.*]] = extractvalue { double, double } [[CALL]], 0
 // CFINITEONLY-NEXT:    [[TMP3:%.*]] = extractvalue { double, double } [[CALL]], 1
 // CFINITEONLY-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // CFINITEONLY:       complex_mul_cont:
 // CFINITEONLY-NEXT:    [[REAL_MUL_PHI:%.*]] = phi nnan ninf double [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP2]], [[COMPLEX_MUL_LIBCALL]] ]
 // CFINITEONLY-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi nnan ninf double [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP3]], [[COMPLEX_MUL_LIBCALL]] ]
 // CFINITEONLY-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store double [[REAL_MUL_PHI]], ptr [[RETVAL_REALP]], align 8
 // CFINITEONLY-NEXT:    store double [[IMAG_MUL_PHI]], ptr [[RETVAL_IMAGP]], align 8
 // CFINITEONLY-NEXT:    [[TMP4:%.*]] = load { double, double }, ptr [[RETVAL]], align 8
 // CFINITEONLY-NEXT:    ret { double, double } [[TMP4]]
 //
 // CLFINITEONLY: Function Attrs: mustprogress nofree norecurse nosync nounwind willreturn memory(none)
 // CLFINITEONLY-LABEL: define dso_local { double, double } @defined_complex_func_f64_ret
-// CLFINITEONLY-SAME: (double noundef [[C_COERCE0:%.*]], double noundef [[C_COERCE1:%.*]]) local_unnamed_addr #[[ATTR0]] {
+// CLFINITEONLY-SAME: (double noundef nofpclass(nan inf) [[C_COERCE0:%.*]], double noundef nofpclass(nan inf) [[C_COERCE1:%.*]]) local_unnamed_addr #[[ATTR0]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[MUL_AD:%.*]] = fmul nnan ninf double [[C_COERCE0]], [[C_COERCE1]]
 // CLFINITEONLY-NEXT:    [[MUL_I:%.*]] = fadd nnan ninf double [[MUL_AD]], [[MUL_AD]]
 // CLFINITEONLY-NEXT:    [[MUL_AC:%.*]] = fmul nnan ninf double [[C_COERCE0]], [[C_COERCE0]]
 // CLFINITEONLY-NEXT:    [[MUL_BD:%.*]] = fmul nnan ninf double [[C_COERCE1]], [[C_COERCE1]]
 // CLFINITEONLY-NEXT:    [[MUL_R:%.*]] = fsub nnan ninf double [[MUL_AC]], [[MUL_BD]]
 // CLFINITEONLY-NEXT:    [[DOTFCA_0_INSERT:%.*]] = insertvalue { double, double } poison, double [[MUL_R]], 0
 // CLFINITEONLY-NEXT:    [[DOTFCA_1_INSERT:%.*]] = insertvalue { double, double } [[DOTFCA_0_INSERT]], double [[MUL_I]], 1
 // CLFINITEONLY-NEXT:    ret { double, double } [[DOTFCA_1_INSERT]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
 // NONANS-LABEL: define dso_local { double, double } @defined_complex_func_f64_ret
-// NONANS-SAME: (double noundef [[C_COERCE0:%.*]], double noundef [[C_COERCE1:%.*]]) #[[ATTR0]] {
+// NONANS-SAME: (double noundef nofpclass(nan) [[C_COERCE0:%.*]], double noundef nofpclass(nan) [[C_COERCE1:%.*]]) #[[ATTR0]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[RETVAL:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[C:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[TMP0:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0
 // NONANS-NEXT:    store double [[C_COERCE0]], ptr [[TMP0]], align 8
 // NONANS-NEXT:    [[TMP1:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 1
 // NONANS-NEXT:    store double [[C_COERCE1]], ptr [[TMP1]], align 8
 // NONANS-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0

 // NONANS-NEXT:    [[MUL_R:%.*]] = fsub nnan double [[MUL_AC]], [[MUL_BD]]
 // NONANS-NEXT:    [[MUL_I:%.*]] = fadd nnan double [[MUL_AD]], [[MUL_BC]]
 // NONANS-NEXT:    [[ISNAN_CMP:%.*]] = fcmp nnan uno double [[MUL_R]], [[MUL_R]]
 // NONANS-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2:![0-9]+]]
 // NONANS:       complex_mul_imag_nan:
 // NONANS-NEXT:    [[ISNAN_CMP5:%.*]] = fcmp nnan uno double [[MUL_I]], [[MUL_I]]
 // NONANS-NEXT:    br i1 [[ISNAN_CMP5]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // NONANS:       complex_mul_libcall:
-// NONANS-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef [[C_REAL]], double noundef [[C_IMAG]], double noundef [[C_REAL2]], double noundef [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
+// NONANS-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef nofpclass(nan) [[C_REAL]], double noundef nofpclass(nan) [[C_IMAG]], double noundef nofpclass(nan) [[C_REAL2]], double noundef nofpclass(nan) [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
 // NONANS-NEXT:    [[TMP2:%.*]] = extractvalue { double, double } [[CALL]], 0
 // NONANS-NEXT:    [[TMP3:%.*]] = extractvalue { double, double } [[CALL]], 1
 // NONANS-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // NONANS:       complex_mul_cont:
 // NONANS-NEXT:    [[REAL_MUL_PHI:%.*]] = phi nnan double [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP2]], [[COMPLEX_MUL_LIBCALL]] ]
 // NONANS-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi nnan double [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP3]], [[COMPLEX_MUL_LIBCALL]] ]
 // NONANS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 0
 // NONANS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 1
 // NONANS-NEXT:    store double [[REAL_MUL_PHI]], ptr [[RETVAL_REALP]], align 8
 // NONANS-NEXT:    store double [[IMAG_MUL_PHI]], ptr [[RETVAL_IMAGP]], align 8
 // NONANS-NEXT:    [[TMP4:%.*]] = load { double, double }, ptr [[RETVAL]], align 8
 // NONANS-NEXT:    ret { double, double } [[TMP4]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
 // NOINFS-LABEL: define dso_local { double, double } @defined_complex_func_f64_ret
-// NOINFS-SAME: (double noundef [[C_COERCE0:%.*]], double noundef [[C_COERCE1:%.*]]) #[[ATTR0]] {
+// NOINFS-SAME: (double noundef nofpclass(inf) [[C_COERCE0:%.*]], double noundef nofpclass(inf) [[C_COERCE1:%.*]]) #[[ATTR0]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[RETVAL:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[C:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[TMP0:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0
 // NOINFS-NEXT:    store double [[C_COERCE0]], ptr [[TMP0]], align 8
 // NOINFS-NEXT:    [[TMP1:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 1
 // NOINFS-NEXT:    store double [[C_COERCE1]], ptr [[TMP1]], align 8
 // NOINFS-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[C]], i32 0, i32 0

 // NOINFS-NEXT:    [[MUL_R:%.*]] = fsub ninf double [[MUL_AC]], [[MUL_BD]]
 // NOINFS-NEXT:    [[MUL_I:%.*]] = fadd ninf double [[MUL_AD]], [[MUL_BC]]
 // NOINFS-NEXT:    [[ISNAN_CMP:%.*]] = fcmp ninf uno double [[MUL_R]], [[MUL_R]]
 // NOINFS-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2:![0-9]+]]
 // NOINFS:       complex_mul_imag_nan:
 // NOINFS-NEXT:    [[ISNAN_CMP5:%.*]] = fcmp ninf uno double [[MUL_I]], [[MUL_I]]
 // NOINFS-NEXT:    br i1 [[ISNAN_CMP5]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // NOINFS:       complex_mul_libcall:
-// NOINFS-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef [[C_REAL]], double noundef [[C_IMAG]], double noundef [[C_REAL2]], double noundef [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
+// NOINFS-NEXT:    [[CALL:%.*]] = call { double, double } @__muldc3(double noundef nofpclass(inf) [[C_REAL]], double noundef nofpclass(inf) [[C_IMAG]], double noundef nofpclass(inf) [[C_REAL2]], double noundef nofpclass(inf) [[C_IMAG4]]) #[[ATTR7:[0-9]+]]
 // NOINFS-NEXT:    [[TMP2:%.*]] = extractvalue { double, double } [[CALL]], 0
 // NOINFS-NEXT:    [[TMP3:%.*]] = extractvalue { double, double } [[CALL]], 1
 // NOINFS-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // NOINFS:       complex_mul_cont:
 // NOINFS-NEXT:    [[REAL_MUL_PHI:%.*]] = phi ninf double [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP2]], [[COMPLEX_MUL_LIBCALL]] ]
 // NOINFS-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi ninf double [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[TMP3]], [[COMPLEX_MUL_LIBCALL]] ]
 // NOINFS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 0
 // NOINFS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[RETVAL]], i32 0, i32 1
 // NOINFS-NEXT:    store double [[REAL_MUL_PHI]], ptr [[RETVAL_REALP]], align 8
 // NOINFS-NEXT:    store double [[IMAG_MUL_PHI]], ptr [[RETVAL_IMAGP]], align 8
 // NOINFS-NEXT:    [[TMP4:%.*]] = load { double, double }, ptr [[RETVAL]], align 8
 // NOINFS-NEXT:    ret { double, double } [[TMP4]]
 //
 _Complex double defined_complex_func_f64_ret(_Complex double c) {
   return c * c;
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local <2 x half> @defined_complex_func_f16_ret
-// CFINITEONLY-SAME: (<2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x half> @defined_complex_func_f16_ret
+// CFINITEONLY-SAME: (<2 x half> noundef nofpclass(nan inf) [[C_COERCE:%.*]]) #[[ATTR4]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[RETVAL:%.*]] = alloca { half, half }, align 2
 // CFINITEONLY-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // CFINITEONLY-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    store <2 x half> [[C_COERCE]], ptr [[C]], align 2
 // CFINITEONLY-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[C_REAL:%.*]] = load half, ptr [[C_REALP]], align 2
 // CFINITEONLY-NEXT:    [[C_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 1

 // CFINITEONLY-NEXT:    [[MUL_R:%.*]] = fsub nnan ninf float [[MUL_AC]], [[MUL_BD]]
 // CFINITEONLY-NEXT:    [[MUL_I:%.*]] = fadd nnan ninf float [[MUL_AD]], [[MUL_BC]]
 // CFINITEONLY-NEXT:    [[ISNAN_CMP:%.*]] = fcmp nnan ninf uno float [[MUL_R]], [[MUL_R]]
 // CFINITEONLY-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2]]
 // CFINITEONLY:       complex_mul_imag_nan:
 // CFINITEONLY-NEXT:    [[ISNAN_CMP8:%.*]] = fcmp nnan ninf uno float [[MUL_I]], [[MUL_I]]
 // CFINITEONLY-NEXT:    br i1 [[ISNAN_CMP8]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // CFINITEONLY:       complex_mul_libcall:
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf <2 x float> @__mulsc3(float noundef [[EXT]], float noundef [[EXT1]], float noundef [[EXT6]], float noundef [[EXT7]]) #[[ATTR7]]
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) <2 x float> @__mulsc3(float noundef nofpclass(nan inf) [[EXT]], float noundef nofpclass(nan inf) [[EXT1]], float noundef nofpclass(nan inf) [[EXT6]], float noundef nofpclass(nan inf) [[EXT7]]) #[[ATTR7]]
 // CFINITEONLY-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE]], align 4
 // CFINITEONLY-NEXT:    [[COERCE_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[COERCE_REAL:%.*]] = load float, ptr [[COERCE_REALP]], align 4
 // CFINITEONLY-NEXT:    [[COERCE_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 1
 // CFINITEONLY-NEXT:    [[COERCE_IMAG:%.*]] = load float, ptr [[COERCE_IMAGP]], align 4
 // CFINITEONLY-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // CFINITEONLY:       complex_mul_cont:
 // CFINITEONLY-NEXT:    [[REAL_MUL_PHI:%.*]] = phi nnan ninf float [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_REAL]], [[COMPLEX_MUL_LIBCALL]] ]
 // CFINITEONLY-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi nnan ninf float [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_IMAG]], [[COMPLEX_MUL_LIBCALL]] ]
 // CFINITEONLY-NEXT:    [[UNPROMOTION:%.*]] = fptrunc float [[REAL_MUL_PHI]] to half
 // CFINITEONLY-NEXT:    [[UNPROMOTION9:%.*]] = fptrunc float [[IMAG_MUL_PHI]] to half
 // CFINITEONLY-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store half [[UNPROMOTION]], ptr [[RETVAL_REALP]], align 2
 // CFINITEONLY-NEXT:    store half [[UNPROMOTION9]], ptr [[RETVAL_IMAGP]], align 2
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load <2 x half>, ptr [[RETVAL]], align 2
 // CFINITEONLY-NEXT:    ret <2 x half> [[TMP0]]
 //
 // CLFINITEONLY: Function Attrs: mustprogress nofree norecurse nosync nounwind willreturn memory(none)
-// CLFINITEONLY-LABEL: define dso_local <2 x half> @defined_complex_func_f16_ret
-// CLFINITEONLY-SAME: (<2 x half> noundef [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR7:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <2 x half> @defined_complex_func_f16_ret
+// CLFINITEONLY-SAME: (<2 x half> noundef nofpclass(nan inf) [[C_COERCE:%.*]]) local_unnamed_addr #[[ATTR7:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[C_SROA_0_0_VEC_EXTRACT:%.*]] = extractelement <2 x half> [[C_COERCE]], i64 0
 // CLFINITEONLY-NEXT:    [[EXT:%.*]] = fpext half [[C_SROA_0_0_VEC_EXTRACT]] to float
 // CLFINITEONLY-NEXT:    [[C_SROA_0_2_VEC_EXTRACT:%.*]] = extractelement <2 x half> [[C_COERCE]], i64 1
 // CLFINITEONLY-NEXT:    [[EXT1:%.*]] = fpext half [[C_SROA_0_2_VEC_EXTRACT]] to float
 // CLFINITEONLY-NEXT:    [[MUL_AD:%.*]] = fmul nnan ninf float [[EXT]], [[EXT1]]
 // CLFINITEONLY-NEXT:    [[MUL_I:%.*]] = fadd nnan ninf float [[MUL_AD]], [[MUL_AD]]
 // CLFINITEONLY-NEXT:    [[MUL_AC:%.*]] = fmul nnan ninf float [[EXT]], [[EXT]]
 // CLFINITEONLY-NEXT:    [[MUL_BD:%.*]] = fmul nnan ninf float [[EXT1]], [[EXT1]]
 // CLFINITEONLY-NEXT:    [[MUL_R:%.*]] = fsub nnan ninf float [[MUL_AC]], [[MUL_BD]]
 // CLFINITEONLY-NEXT:    [[UNPROMOTION:%.*]] = fptrunc float [[MUL_R]] to half
 // CLFINITEONLY-NEXT:    [[UNPROMOTION9:%.*]] = fptrunc float [[MUL_I]] to half
 // CLFINITEONLY-NEXT:    [[RETVAL_SROA_0_0_VEC_INSERT:%.*]] = insertelement <2 x half> undef, half [[UNPROMOTION]], i64 0
 // CLFINITEONLY-NEXT:    [[RETVAL_SROA_0_2_VEC_INSERT:%.*]] = insertelement <2 x half> [[RETVAL_SROA_0_0_VEC_INSERT]], half [[UNPROMOTION9]], i64 1
 // CLFINITEONLY-NEXT:    ret <2 x half> [[RETVAL_SROA_0_2_VEC_INSERT]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local <2 x half> @defined_complex_func_f16_ret
-// NONANS-SAME: (<2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) <2 x half> @defined_complex_func_f16_ret
+// NONANS-SAME: (<2 x half> noundef nofpclass(nan) [[C_COERCE:%.*]]) #[[ATTR4]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[RETVAL:%.*]] = alloca { half, half }, align 2
 // NONANS-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // NONANS-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    store <2 x half> [[C_COERCE]], ptr [[C]], align 2
 // NONANS-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 0
 // NONANS-NEXT:    [[C_REAL:%.*]] = load half, ptr [[C_REALP]], align 2
 // NONANS-NEXT:    [[C_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 1

 // NONANS-NEXT:    [[MUL_R:%.*]] = fsub nnan float [[MUL_AC]], [[MUL_BD]]
 // NONANS-NEXT:    [[MUL_I:%.*]] = fadd nnan float [[MUL_AD]], [[MUL_BC]]
 // NONANS-NEXT:    [[ISNAN_CMP:%.*]] = fcmp nnan uno float [[MUL_R]], [[MUL_R]]
 // NONANS-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2]]
 // NONANS:       complex_mul_imag_nan:
 // NONANS-NEXT:    [[ISNAN_CMP8:%.*]] = fcmp nnan uno float [[MUL_I]], [[MUL_I]]
 // NONANS-NEXT:    br i1 [[ISNAN_CMP8]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // NONANS:       complex_mul_libcall:
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan <2 x float> @__mulsc3(float noundef [[EXT]], float noundef [[EXT1]], float noundef [[EXT6]], float noundef [[EXT7]]) #[[ATTR7]]
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) <2 x float> @__mulsc3(float noundef nofpclass(nan) [[EXT]], float noundef nofpclass(nan) [[EXT1]], float noundef nofpclass(nan) [[EXT6]], float noundef nofpclass(nan) [[EXT7]]) #[[ATTR7]]
 // NONANS-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE]], align 4
 // NONANS-NEXT:    [[COERCE_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 0
 // NONANS-NEXT:    [[COERCE_REAL:%.*]] = load float, ptr [[COERCE_REALP]], align 4
 // NONANS-NEXT:    [[COERCE_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 1
 // NONANS-NEXT:    [[COERCE_IMAG:%.*]] = load float, ptr [[COERCE_IMAGP]], align 4
 // NONANS-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // NONANS:       complex_mul_cont:
 // NONANS-NEXT:    [[REAL_MUL_PHI:%.*]] = phi nnan float [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_REAL]], [[COMPLEX_MUL_LIBCALL]] ]
 // NONANS-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi nnan float [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_IMAG]], [[COMPLEX_MUL_LIBCALL]] ]
 // NONANS-NEXT:    [[UNPROMOTION:%.*]] = fptrunc float [[REAL_MUL_PHI]] to half
 // NONANS-NEXT:    [[UNPROMOTION9:%.*]] = fptrunc float [[IMAG_MUL_PHI]] to half
 // NONANS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 0
 // NONANS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 1
 // NONANS-NEXT:    store half [[UNPROMOTION]], ptr [[RETVAL_REALP]], align 2
 // NONANS-NEXT:    store half [[UNPROMOTION9]], ptr [[RETVAL_IMAGP]], align 2
 // NONANS-NEXT:    [[TMP0:%.*]] = load <2 x half>, ptr [[RETVAL]], align 2
 // NONANS-NEXT:    ret <2 x half> [[TMP0]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local <2 x half> @defined_complex_func_f16_ret
-// NOINFS-SAME: (<2 x half> noundef [[C_COERCE:%.*]]) #[[ATTR4]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) <2 x half> @defined_complex_func_f16_ret
+// NOINFS-SAME: (<2 x half> noundef nofpclass(inf) [[C_COERCE:%.*]]) #[[ATTR4]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[RETVAL:%.*]] = alloca { half, half }, align 2
 // NOINFS-NEXT:    [[C:%.*]] = alloca { half, half }, align 2
 // NOINFS-NEXT:    [[COERCE:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    store <2 x half> [[C_COERCE]], ptr [[C]], align 2
 // NOINFS-NEXT:    [[C_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 0
 // NOINFS-NEXT:    [[C_REAL:%.*]] = load half, ptr [[C_REALP]], align 2
 // NOINFS-NEXT:    [[C_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[C]], i32 0, i32 1

 // NOINFS-NEXT:    [[MUL_R:%.*]] = fsub ninf float [[MUL_AC]], [[MUL_BD]]
 // NOINFS-NEXT:    [[MUL_I:%.*]] = fadd ninf float [[MUL_AD]], [[MUL_BC]]
 // NOINFS-NEXT:    [[ISNAN_CMP:%.*]] = fcmp ninf uno float [[MUL_R]], [[MUL_R]]
 // NOINFS-NEXT:    br i1 [[ISNAN_CMP]], label [[COMPLEX_MUL_IMAG_NAN:%.*]], label [[COMPLEX_MUL_CONT:%.*]], !prof [[PROF2]]
 // NOINFS:       complex_mul_imag_nan:
 // NOINFS-NEXT:    [[ISNAN_CMP8:%.*]] = fcmp ninf uno float [[MUL_I]], [[MUL_I]]
 // NOINFS-NEXT:    br i1 [[ISNAN_CMP8]], label [[COMPLEX_MUL_LIBCALL:%.*]], label [[COMPLEX_MUL_CONT]], !prof [[PROF2]]
 // NOINFS:       complex_mul_libcall:
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf <2 x float> @__mulsc3(float noundef [[EXT]], float noundef [[EXT1]], float noundef [[EXT6]], float noundef [[EXT7]]) #[[ATTR7]]
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) <2 x float> @__mulsc3(float noundef nofpclass(inf) [[EXT]], float noundef nofpclass(inf) [[EXT1]], float noundef nofpclass(inf) [[EXT6]], float noundef nofpclass(inf) [[EXT7]]) #[[ATTR7]]
 // NOINFS-NEXT:    store <2 x float> [[CALL]], ptr [[COERCE]], align 4
 // NOINFS-NEXT:    [[COERCE_REALP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 0
 // NOINFS-NEXT:    [[COERCE_REAL:%.*]] = load float, ptr [[COERCE_REALP]], align 4
 // NOINFS-NEXT:    [[COERCE_IMAGP:%.*]] = getelementptr inbounds { float, float }, ptr [[COERCE]], i32 0, i32 1
 // NOINFS-NEXT:    [[COERCE_IMAG:%.*]] = load float, ptr [[COERCE_IMAGP]], align 4
 // NOINFS-NEXT:    br label [[COMPLEX_MUL_CONT]]
 // NOINFS:       complex_mul_cont:
 // NOINFS-NEXT:    [[REAL_MUL_PHI:%.*]] = phi ninf float [ [[MUL_R]], [[ENTRY:%.*]] ], [ [[MUL_R]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_REAL]], [[COMPLEX_MUL_LIBCALL]] ]
 // NOINFS-NEXT:    [[IMAG_MUL_PHI:%.*]] = phi ninf float [ [[MUL_I]], [[ENTRY]] ], [ [[MUL_I]], [[COMPLEX_MUL_IMAG_NAN]] ], [ [[COERCE_IMAG]], [[COMPLEX_MUL_LIBCALL]] ]
 // NOINFS-NEXT:    [[UNPROMOTION:%.*]] = fptrunc float [[REAL_MUL_PHI]] to half
 // NOINFS-NEXT:    [[UNPROMOTION9:%.*]] = fptrunc float [[IMAG_MUL_PHI]] to half
 // NOINFS-NEXT:    [[RETVAL_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 0
 // NOINFS-NEXT:    [[RETVAL_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[RETVAL]], i32 0, i32 1
 // NOINFS-NEXT:    store half [[UNPROMOTION]], ptr [[RETVAL_REALP]], align 2
 // NOINFS-NEXT:    store half [[UNPROMOTION9]], ptr [[RETVAL_IMAGP]], align 2
 // NOINFS-NEXT:    [[TMP0:%.*]] = load <2 x half>, ptr [[RETVAL]], align 2
 // NOINFS-NEXT:    ret <2 x half> [[TMP0]]
 //
 _Complex _Float16 defined_complex_func_f16_ret(_Complex _Float16 c) {
   return c * c;
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local float @call_variadic
-// CFINITEONLY-SAME: (float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_variadic
+// CFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[F32:%.*]], double noundef nofpclass(nan inf) [[F64:%.*]], half noundef nofpclass(nan inf) [[F16:%.*]], double noundef nofpclass(nan inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // CFINITEONLY-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8
 // CFINITEONLY-NEXT:    [[F16_ADDR:%.*]] = alloca half, align 2

 // CFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // CFINITEONLY-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // CFINITEONLY-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // CFINITEONLY-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // CFINITEONLY-NEXT:    [[TMP12:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf float (float, ...) @variadic(float noundef [[TMP2]], double noundef [[CONV]], double noundef [[TMP4]], half noundef [[TMP5]], double noundef [[TMP9]], <2 x double> noundef [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP12]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) float (float, ...) @variadic(float noundef nofpclass(nan inf) [[TMP2]], double noundef nofpclass(nan inf) [[CONV]], double noundef nofpclass(nan inf) [[TMP4]], half noundef nofpclass(nan inf) [[TMP5]], double noundef nofpclass(nan inf) [[TMP9]], <2 x double> noundef nofpclass(nan inf) [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef nofpclass(nan inf) [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan inf) [[TMP12]])
 // CFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local float @call_variadic
-// CLFINITEONLY-SAME: (float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr nocapture noundef readonly byval({ half, half }) align 8 [[CF16:%.*]]) local_unnamed_addr #[[ATTR5]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_variadic
+// CLFINITEONLY-SAME: (float noundef nofpclass(nan inf) [[F32:%.*]], double noundef nofpclass(nan inf) [[F64:%.*]], half noundef nofpclass(nan inf) [[F16:%.*]], double noundef nofpclass(nan inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE1:%.*]], ptr nocapture noundef readonly byval({ half, half }) align 8 [[CF16:%.*]]) local_unnamed_addr #[[ATTR5]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP:%.*]] = alloca { double, double }, align 8
 // CLFINITEONLY-NEXT:    [[CONV:%.*]] = fpext float [[F32]] to double
 // CLFINITEONLY-NEXT:    [[CF16_REAL:%.*]] = load half, ptr [[CF16]], align 8
 // CLFINITEONLY-NEXT:    [[CF16_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[CF16]], i64 0, i32 1
 // CLFINITEONLY-NEXT:    [[CF16_IMAG:%.*]] = load half, ptr [[CF16_IMAGP]], align 2
 // CLFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i64 0, i32 1
 // CLFINITEONLY-NEXT:    store double [[CF64_COERCE0]], ptr [[INDIRECT_ARG_TEMP]], align 8
 // CLFINITEONLY-NEXT:    store double [[CF64_COERCE1]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // CLFINITEONLY-NEXT:    [[COERCE5_SROA_0_0_VEC_INSERT:%.*]] = insertelement <2 x half> undef, half [[CF16_REAL]], i64 0
 // CLFINITEONLY-NEXT:    [[COERCE5_SROA_0_2_VEC_INSERT:%.*]] = insertelement <2 x half> [[COERCE5_SROA_0_0_VEC_INSERT]], half [[CF16_IMAG]], i64 1
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf float (float, ...) @variadic(float noundef [[F32]], double noundef [[CONV]], double noundef [[F64]], half noundef [[F16]], double noundef [[V2F32_COERCE]], <2 x double> noundef [[V2F64]], i32 noundef [[V2F16_COERCE]], <2 x float> noundef [[CF32_COERCE]], ptr noundef nonnull byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[COERCE5_SROA_0_2_VEC_INSERT]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) float (float, ...) @variadic(float noundef nofpclass(nan inf) [[F32]], double noundef nofpclass(nan inf) [[CONV]], double noundef nofpclass(nan inf) [[F64]], half noundef nofpclass(nan inf) [[F16]], double noundef nofpclass(nan inf) [[V2F32_COERCE]], <2 x double> noundef nofpclass(nan inf) [[V2F64]], i32 noundef [[V2F16_COERCE]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE]], ptr noundef nonnull byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan inf) [[COERCE5_SROA_0_2_VEC_INSERT]]) #[[ATTR10]]
 // CLFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local float @call_variadic
-// NONANS-SAME: (float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) float @call_variadic
+// NONANS-SAME: (float noundef nofpclass(nan) [[F32:%.*]], double noundef nofpclass(nan) [[F64:%.*]], half noundef nofpclass(nan) [[F16:%.*]], double noundef nofpclass(nan) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan) [[CF32_COERCE:%.*]], double noundef nofpclass(nan) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // NONANS-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8
 // NONANS-NEXT:    [[F16_ADDR:%.*]] = alloca half, align 2

 // NONANS-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // NONANS-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // NONANS-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // NONANS-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // NONANS-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // NONANS-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // NONANS-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // NONANS-NEXT:    [[TMP12:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan float (float, ...) @variadic(float noundef [[TMP2]], double noundef [[CONV]], double noundef [[TMP4]], half noundef [[TMP5]], double noundef [[TMP9]], <2 x double> noundef [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP12]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) float (float, ...) @variadic(float noundef nofpclass(nan) [[TMP2]], double noundef nofpclass(nan) [[CONV]], double noundef nofpclass(nan) [[TMP4]], half noundef nofpclass(nan) [[TMP5]], double noundef nofpclass(nan) [[TMP9]], <2 x double> noundef nofpclass(nan) [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef nofpclass(nan) [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan) [[TMP12]])
 // NONANS-NEXT:    ret float [[CALL]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local float @call_variadic
-// NOINFS-SAME: (float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) float @call_variadic
+// NOINFS-SAME: (float noundef nofpclass(inf) [[F32:%.*]], double noundef nofpclass(inf) [[F64:%.*]], half noundef nofpclass(inf) [[F16:%.*]], double noundef nofpclass(inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(inf) [[CF32_COERCE:%.*]], double noundef nofpclass(inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(inf) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // NOINFS-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8
 // NOINFS-NEXT:    [[F16_ADDR:%.*]] = alloca half, align 2

 // NOINFS-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // NOINFS-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // NOINFS-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // NOINFS-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // NOINFS-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // NOINFS-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // NOINFS-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // NOINFS-NEXT:    [[TMP12:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf float (float, ...) @variadic(float noundef [[TMP2]], double noundef [[CONV]], double noundef [[TMP4]], half noundef [[TMP5]], double noundef [[TMP9]], <2 x double> noundef [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP12]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) float (float, ...) @variadic(float noundef nofpclass(inf) [[TMP2]], double noundef nofpclass(inf) [[CONV]], double noundef nofpclass(inf) [[TMP4]], half noundef nofpclass(inf) [[TMP5]], double noundef nofpclass(inf) [[TMP9]], <2 x double> noundef nofpclass(inf) [[TMP7]], i32 noundef [[TMP10]], <2 x float> noundef nofpclass(inf) [[TMP11]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(inf) [[TMP12]])
 // NOINFS-NEXT:    ret float [[CALL]]
 //
 float call_variadic(float f32, double f64, _Float16 f16,
                     float2 v2f32, double2 v2f64, half2 v2f16,
                     _Complex float cf32, _Complex double cf64, _Complex _Float16 cf16) {
   return variadic(f32, f32, f64, f16, v2f32, v2f64, v2f16, cf32, cf64, cf16);
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local float @call_variadic_indirect
-// CFINITEONLY-SAME: (ptr noundef [[FPTR:%.*]], float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_variadic_indirect
+// CFINITEONLY-SAME: (ptr noundef [[FPTR:%.*]], float noundef nofpclass(nan inf) [[F32:%.*]], double noundef nofpclass(nan inf) [[F64:%.*]], half noundef nofpclass(nan inf) [[F16:%.*]], double noundef nofpclass(nan inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // CFINITEONLY-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // CFINITEONLY-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // CFINITEONLY-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // CFINITEONLY-NEXT:    [[FPTR_ADDR:%.*]] = alloca ptr, align 8
 // CFINITEONLY-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // CFINITEONLY-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8

 // CFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // CFINITEONLY-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // CFINITEONLY-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // CFINITEONLY-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // CFINITEONLY-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // CFINITEONLY-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // CFINITEONLY-NEXT:    [[TMP13:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf float (float, ...) [[TMP2]](float noundef [[TMP3]], double noundef [[CONV]], double noundef [[TMP5]], half noundef [[TMP6]], double noundef [[TMP10]], <2 x double> noundef [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP13]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) float (float, ...) [[TMP2]](float noundef nofpclass(nan inf) [[TMP3]], double noundef nofpclass(nan inf) [[CONV]], double noundef nofpclass(nan inf) [[TMP5]], half noundef nofpclass(nan inf) [[TMP6]], double noundef nofpclass(nan inf) [[TMP10]], <2 x double> noundef nofpclass(nan inf) [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef nofpclass(nan inf) [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan inf) [[TMP13]])
 // CFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local float @call_variadic_indirect
-// CLFINITEONLY-SAME: (ptr nocapture noundef readonly [[FPTR:%.*]], float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr nocapture noundef readonly byval({ half, half }) align 8 [[CF16:%.*]]) local_unnamed_addr #[[ATTR5]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) float @call_variadic_indirect
+// CLFINITEONLY-SAME: (ptr nocapture noundef readonly [[FPTR:%.*]], float noundef nofpclass(nan inf) [[F32:%.*]], double noundef nofpclass(nan inf) [[F64:%.*]], half noundef nofpclass(nan inf) [[F16:%.*]], double noundef nofpclass(nan inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan inf) [[CF64_COERCE1:%.*]], ptr nocapture noundef readonly byval({ half, half }) align 8 [[CF16:%.*]]) local_unnamed_addr #[[ATTR5]] {
 // CLFINITEONLY-NEXT:  entry:
 // CLFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP:%.*]] = alloca { double, double }, align 8
 // CLFINITEONLY-NEXT:    [[CONV:%.*]] = fpext float [[F32]] to double
 // CLFINITEONLY-NEXT:    [[CF16_REAL:%.*]] = load half, ptr [[CF16]], align 8
 // CLFINITEONLY-NEXT:    [[CF16_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[CF16]], i64 0, i32 1
 // CLFINITEONLY-NEXT:    [[CF16_IMAG:%.*]] = load half, ptr [[CF16_IMAGP]], align 2
 // CLFINITEONLY-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i64 0, i32 1
 // CLFINITEONLY-NEXT:    store double [[CF64_COERCE0]], ptr [[INDIRECT_ARG_TEMP]], align 8
 // CLFINITEONLY-NEXT:    store double [[CF64_COERCE1]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // CLFINITEONLY-NEXT:    [[COERCE5_SROA_0_0_VEC_INSERT:%.*]] = insertelement <2 x half> undef, half [[CF16_REAL]], i64 0
 // CLFINITEONLY-NEXT:    [[COERCE5_SROA_0_2_VEC_INSERT:%.*]] = insertelement <2 x half> [[COERCE5_SROA_0_0_VEC_INSERT]], half [[CF16_IMAG]], i64 1
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf float (float, ...) [[FPTR]](float noundef [[F32]], double noundef [[CONV]], double noundef [[F64]], half noundef [[F16]], double noundef [[V2F32_COERCE]], <2 x double> noundef [[V2F64]], i32 noundef [[V2F16_COERCE]], <2 x float> noundef [[CF32_COERCE]], ptr noundef nonnull byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[COERCE5_SROA_0_2_VEC_INSERT]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) float (float, ...) [[FPTR]](float noundef nofpclass(nan inf) [[F32]], double noundef nofpclass(nan inf) [[CONV]], double noundef nofpclass(nan inf) [[F64]], half noundef nofpclass(nan inf) [[F16]], double noundef nofpclass(nan inf) [[V2F32_COERCE]], <2 x double> noundef nofpclass(nan inf) [[V2F64]], i32 noundef [[V2F16_COERCE]], <2 x float> noundef nofpclass(nan inf) [[CF32_COERCE]], ptr noundef nonnull byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan inf) [[COERCE5_SROA_0_2_VEC_INSERT]]) #[[ATTR10]]
 // CLFINITEONLY-NEXT:    ret float [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local float @call_variadic_indirect
-// NONANS-SAME: (ptr noundef [[FPTR:%.*]], float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) float @call_variadic_indirect
+// NONANS-SAME: (ptr noundef [[FPTR:%.*]], float noundef nofpclass(nan) [[F32:%.*]], double noundef nofpclass(nan) [[F64:%.*]], half noundef nofpclass(nan) [[F16:%.*]], double noundef nofpclass(nan) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(nan) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(nan) [[CF32_COERCE:%.*]], double noundef nofpclass(nan) [[CF64_COERCE0:%.*]], double noundef nofpclass(nan) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // NONANS-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // NONANS-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // NONANS-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // NONANS-NEXT:    [[FPTR_ADDR:%.*]] = alloca ptr, align 8
 // NONANS-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // NONANS-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8

 // NONANS-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // NONANS-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // NONANS-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // NONANS-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // NONANS-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // NONANS-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // NONANS-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // NONANS-NEXT:    [[TMP13:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan float (float, ...) [[TMP2]](float noundef [[TMP3]], double noundef [[CONV]], double noundef [[TMP5]], half noundef [[TMP6]], double noundef [[TMP10]], <2 x double> noundef [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP13]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) float (float, ...) [[TMP2]](float noundef nofpclass(nan) [[TMP3]], double noundef nofpclass(nan) [[CONV]], double noundef nofpclass(nan) [[TMP5]], half noundef nofpclass(nan) [[TMP6]], double noundef nofpclass(nan) [[TMP10]], <2 x double> noundef nofpclass(nan) [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef nofpclass(nan) [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(nan) [[TMP13]])
 // NONANS-NEXT:    ret float [[CALL]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local float @call_variadic_indirect
-// NOINFS-SAME: (ptr noundef [[FPTR:%.*]], float noundef [[F32:%.*]], double noundef [[F64:%.*]], half noundef [[F16:%.*]], double noundef [[V2F32_COERCE:%.*]], <2 x double> noundef [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef [[CF32_COERCE:%.*]], double noundef [[CF64_COERCE0:%.*]], double noundef [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) float @call_variadic_indirect
+// NOINFS-SAME: (ptr noundef [[FPTR:%.*]], float noundef nofpclass(inf) [[F32:%.*]], double noundef nofpclass(inf) [[F64:%.*]], half noundef nofpclass(inf) [[F16:%.*]], double noundef nofpclass(inf) [[V2F32_COERCE:%.*]], <2 x double> noundef nofpclass(inf) [[V2F64:%.*]], i32 noundef [[V2F16_COERCE:%.*]], <2 x float> noundef nofpclass(inf) [[CF32_COERCE:%.*]], double noundef nofpclass(inf) [[CF64_COERCE0:%.*]], double noundef nofpclass(inf) [[CF64_COERCE1:%.*]], ptr noundef byval({ half, half }) align 8 [[CF16:%.*]]) #[[ATTR2]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[V2F32:%.*]] = alloca <2 x float>, align 8
 // NOINFS-NEXT:    [[V2F16:%.*]] = alloca <2 x half>, align 4
 // NOINFS-NEXT:    [[CF32:%.*]] = alloca { float, float }, align 4
 // NOINFS-NEXT:    [[CF64:%.*]] = alloca { double, double }, align 8
 // NOINFS-NEXT:    [[FPTR_ADDR:%.*]] = alloca ptr, align 8
 // NOINFS-NEXT:    [[F32_ADDR:%.*]] = alloca float, align 4
 // NOINFS-NEXT:    [[F64_ADDR:%.*]] = alloca double, align 8

 // NOINFS-NEXT:    [[INDIRECT_ARG_TEMP_IMAGP:%.*]] = getelementptr inbounds { double, double }, ptr [[INDIRECT_ARG_TEMP]], i32 0, i32 1
 // NOINFS-NEXT:    store double [[CF64_REAL]], ptr [[INDIRECT_ARG_TEMP_REALP]], align 8
 // NOINFS-NEXT:    store double [[CF64_IMAG]], ptr [[INDIRECT_ARG_TEMP_IMAGP]], align 8
 // NOINFS-NEXT:    [[COERCE5_REALP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 0
 // NOINFS-NEXT:    [[COERCE5_IMAGP:%.*]] = getelementptr inbounds { half, half }, ptr [[COERCE5]], i32 0, i32 1
 // NOINFS-NEXT:    store half [[CF16_REAL]], ptr [[COERCE5_REALP]], align 2
 // NOINFS-NEXT:    store half [[CF16_IMAG]], ptr [[COERCE5_IMAGP]], align 2
 // NOINFS-NEXT:    [[TMP13:%.*]] = load <2 x half>, ptr [[COERCE5]], align 2
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf float (float, ...) [[TMP2]](float noundef [[TMP3]], double noundef [[CONV]], double noundef [[TMP5]], half noundef [[TMP6]], double noundef [[TMP10]], <2 x double> noundef [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef [[TMP13]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) float (float, ...) [[TMP2]](float noundef nofpclass(inf) [[TMP3]], double noundef nofpclass(inf) [[CONV]], double noundef nofpclass(inf) [[TMP5]], half noundef nofpclass(inf) [[TMP6]], double noundef nofpclass(inf) [[TMP10]], <2 x double> noundef nofpclass(inf) [[TMP8]], i32 noundef [[TMP11]], <2 x float> noundef nofpclass(inf) [[TMP12]], ptr noundef byval({ double, double }) align 8 [[INDIRECT_ARG_TEMP]], <2 x half> noundef nofpclass(inf) [[TMP13]])
 // NOINFS-NEXT:    ret float [[CALL]]
 //
 float call_variadic_indirect(float fptr(float, ...), float f32, double f64, _Float16 f16,
                              float2 v2f32, double2 v2f64, half2 v2f16,
                              _Complex float cf32, _Complex double cf64, _Complex _Float16 cf16) {
   return (*fptr)(f32, f32, f64, f16, v2f32, v2f64, v2f16, cf32, cf64, cf16);
 }
 
 typedef __attribute__((__vector_size__(4 * sizeof(double)))) double __m256d;
 extern __m256d extern_m256d(__m256d, ...);
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
-// CFINITEONLY-LABEL: define dso_local <4 x double> @call_m256d
-// CFINITEONLY-SAME: (<4 x double> noundef [[X:%.*]]) #[[ATTR5:[0-9]+]] {
+// CFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <4 x double> @call_m256d
+// CFINITEONLY-SAME: (<4 x double> noundef nofpclass(nan inf) [[X:%.*]]) #[[ATTR5:[0-9]+]] {
 // CFINITEONLY-NEXT:  entry:
 // CFINITEONLY-NEXT:    [[X_ADDR:%.*]] = alloca <4 x double>, align 32
 // CFINITEONLY-NEXT:    store <4 x double> [[X]], ptr [[X_ADDR]], align 32
 // CFINITEONLY-NEXT:    [[TMP0:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
 // CFINITEONLY-NEXT:    [[TMP1:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
-// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef [[TMP0]], <4 x double> noundef [[TMP1]])
+// CFINITEONLY-NEXT:    [[CALL:%.*]] = call nnan ninf nofpclass(nan inf) <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef nofpclass(nan inf) [[TMP0]], <4 x double> noundef nofpclass(nan inf) [[TMP1]])
 // CFINITEONLY-NEXT:    ret <4 x double> [[CALL]]
 //
 // CLFINITEONLY: Function Attrs: convergent norecurse nounwind
-// CLFINITEONLY-LABEL: define dso_local <4 x double> @call_m256d
-// CLFINITEONLY-SAME: (<4 x double> noundef [[X:%.*]]) local_unnamed_addr #[[ATTR8:[0-9]+]] {
+// CLFINITEONLY-LABEL: define dso_local nofpclass(nan inf) <4 x double> @call_m256d
+// CLFINITEONLY-SAME: (<4 x double> noundef nofpclass(nan inf) [[X:%.*]]) local_unnamed_addr #[[ATTR8:[0-9]+]] {
 // CLFINITEONLY-NEXT:  entry:
-// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef [[X]], <4 x double> noundef [[X]]) #[[ATTR10]]
+// CLFINITEONLY-NEXT:    [[CALL:%.*]] = tail call nnan ninf nofpclass(nan inf) <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef nofpclass(nan inf) [[X]], <4 x double> noundef nofpclass(nan inf) [[X]]) #[[ATTR10]]
 // CLFINITEONLY-NEXT:    ret <4 x double> [[CALL]]
 //
 // NONANS: Function Attrs: noinline nounwind optnone
-// NONANS-LABEL: define dso_local <4 x double> @call_m256d
-// NONANS-SAME: (<4 x double> noundef [[X:%.*]]) #[[ATTR5:[0-9]+]] {
+// NONANS-LABEL: define dso_local nofpclass(nan) <4 x double> @call_m256d
+// NONANS-SAME: (<4 x double> noundef nofpclass(nan) [[X:%.*]]) #[[ATTR5:[0-9]+]] {
 // NONANS-NEXT:  entry:
 // NONANS-NEXT:    [[X_ADDR:%.*]] = alloca <4 x double>, align 32
 // NONANS-NEXT:    store <4 x double> [[X]], ptr [[X_ADDR]], align 32
 // NONANS-NEXT:    [[TMP0:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
 // NONANS-NEXT:    [[TMP1:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
-// NONANS-NEXT:    [[CALL:%.*]] = call nnan <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef [[TMP0]], <4 x double> noundef [[TMP1]])
+// NONANS-NEXT:    [[CALL:%.*]] = call nnan nofpclass(nan) <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef nofpclass(nan) [[TMP0]], <4 x double> noundef nofpclass(nan) [[TMP1]])
 // NONANS-NEXT:    ret <4 x double> [[CALL]]
 //
 // NOINFS: Function Attrs: noinline nounwind optnone
-// NOINFS-LABEL: define dso_local <4 x double> @call_m256d
-// NOINFS-SAME: (<4 x double> noundef [[X:%.*]]) #[[ATTR5:[0-9]+]] {
+// NOINFS-LABEL: define dso_local nofpclass(inf) <4 x double> @call_m256d
+// NOINFS-SAME: (<4 x double> noundef nofpclass(inf) [[X:%.*]]) #[[ATTR5:[0-9]+]] {
 // NOINFS-NEXT:  entry:
 // NOINFS-NEXT:    [[X_ADDR:%.*]] = alloca <4 x double>, align 32
 // NOINFS-NEXT:    store <4 x double> [[X]], ptr [[X_ADDR]], align 32
 // NOINFS-NEXT:    [[TMP0:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
 // NOINFS-NEXT:    [[TMP1:%.*]] = load <4 x double>, ptr [[X_ADDR]], align 32
-// NOINFS-NEXT:    [[CALL:%.*]] = call ninf <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef [[TMP0]], <4 x double> noundef [[TMP1]])
+// NOINFS-NEXT:    [[CALL:%.*]] = call ninf nofpclass(inf) <4 x double> (<4 x double>, ...) @extern_m256d(<4 x double> noundef nofpclass(inf) [[TMP0]], <4 x double> noundef nofpclass(inf) [[TMP1]])
 // NOINFS-NEXT:    ret <4 x double> [[CALL]]
 //
 __m256d call_m256d(__m256d x) {
   return extern_m256d(x, x);
 }
 
 // CFINITEONLY: Function Attrs: noinline nounwind optnone
 // CFINITEONLY-LABEL: define dso_local <25 x double> @call_matrix

llvm/include/llvm/IR/Attributes.h

 namespace AttributeFuncs {
 
 enum AttributeSafetyKind : uint8_t {
   ASK_SAFE_TO_DROP = 1,
   ASK_UNSAFE_TO_DROP = 2,
   ASK_ALL = ASK_SAFE_TO_DROP | ASK_UNSAFE_TO_DROP,
 };
 
+/// Returns true if this is a type legal for the 'nofpclass' attribute. This
+/// follows the same type rules as FPMathOperator.
+bool isNoFPClassCompatibleType(Type *Ty);
+
 /// Which attributes cannot be applied to a type. The argument \p ASK indicates,
 /// if only attributes that are known to be safely droppable are contained in
 /// the mask; only attributes that might be unsafe to drop (e.g., ABI-related
 /// attributes) are in the mask; or both.
 AttributeMask typeIncompatible(Type *Ty, AttributeSafetyKind ASK = ASK_ALL);
 
 /// Get param/return attributes which imply immediate undefined behavior if an
 /// invalid value is passed. For example, this includes noundef (where undef

llvm/lib/IR/Attributes.cpp

 bool Attribute::isExistingAttribute(StringRef Name) {
   return StringSwitch<bool>(Name)
 #define GET_ATTR_NAMES
 #define ATTRIBUTE_ALL(ENUM_NAME, DISPLAY_NAME) .Case(#DISPLAY_NAME, true)
 #include "llvm/IR/Attributes.inc"
       .Default(false);
 }
 
-/// Returns true if this is a type legal for the 'nofpclass' attribute. This
-/// follows the same type rules as FPMathOperator.
-///
-/// TODO: Consider relaxing to any FP type struct fields.
-static bool isNoFPClassCompatibleType(Type *Ty) {
-  while (ArrayType *ArrTy = dyn_cast<ArrayType>(Ty))
-    Ty = ArrTy->getElementType();
-  return Ty->isFPOrFPVectorTy();
-}
-
 //===----------------------------------------------------------------------===//
 // Attribute Accessor Methods
 //===----------------------------------------------------------------------===//
 
 bool Attribute::isEnumAttribute() const {
   return pImpl && pImpl->isEnumAttribute();
 }
 
AttrBuilder &AttrBuilder::addUWTableAttr(UWTableKind Kind) {
   return addRawIntAttr(Attribute::UWTable, uint64_t(Kind));
 }
 
 AttrBuilder &AttrBuilder::addMemoryAttr(MemoryEffects ME) {
   return addRawIntAttr(Attribute::Memory, ME.toIntValue());
 }
 
 AttrBuilder &AttrBuilder::addNoFPClassAttr(FPClassTest Mask) {
+  if (Mask == fcNone)
+    return *this;
+
   return addRawIntAttr(Attribute::NoFPClass, Mask);
 }
 
 AttrBuilder &AttrBuilder::addAllocKindAttr(AllocFnKind Kind) {
   return addRawIntAttr(Attribute::AllocKind, static_cast<uint64_t>(Kind));
 }
 
 Type *AttrBuilder::getTypeAttr(Attribute::AttrKind Kind) const {

 bool AttrBuilder::operator==(const AttrBuilder &B) const {
   return Attrs == B.Attrs;
 }
 
 //===----------------------------------------------------------------------===//
 // AttributeFuncs Function Defintions
 //===----------------------------------------------------------------------===//
 
+/// Returns true if this is a type legal for the 'nofpclass' attribute. This
+/// follows the same type rules as FPMathOperator.
+///
+/// TODO: Consider relaxing to any FP type struct fields.
+bool AttributeFuncs::isNoFPClassCompatibleType(Type *Ty) {
+  while (ArrayType *ArrTy = dyn_cast<ArrayType>(Ty))
+    Ty = ArrTy->getElementType();
+  return Ty->isFPOrFPVectorTy();
+}
+
 /// Which attributes cannot be applied to a type.
 AttributeMask AttributeFuncs::typeIncompatible(Type *Ty,
                                                AttributeSafetyKind ASK) {
   AttributeMask Incompatible;
 
   if (!Ty->isIntegerTy()) {
     // Attributes that only apply to integers.
     if (ASK & ASK_SAFE_TO_DROP)

llvm/test/Linker/Inputs/nofpclass.ll

+define float @declared_as_nonan(float %arg) {
+  %add = fadd float %arg, 1.0
+  ret float %add
+}

llvm/test/Linker/nofpclass.ll

+; RUN: llvm-link %s %S/Inputs/nofpclass.ll -S -o - | FileCheck -check-prefix=ORDER1 %s
+; RUN: llvm-link %S/Inputs/nofpclass.ll %s -S -o - | FileCheck -check-prefix=ORDER2 %s
+
+; Make sure nofpclass is dropped if the function was declared as
+; nofpclass, but not defined with nofpclass.
+
+; ORDER1: define float @caller(float %arg) {
+; ORDER1-NEXT: %result = call float @declared_as_nonan(float %arg)
+; ORDER1-NEXT: ret float %result
+
+; ORDER1: define float @declared_as_nonan(float %arg) {
+; ORDER1-NEXT: %add = fadd float %arg, 1.000000e+00
+; ORDER1-NEXT: ret float %add
+
+
+; ORDER2: define float @declared_as_nonan(float %arg) {
+; ORDER2-NEXT: %add = fadd float %arg, 1.000000e+00
+; ORDER2-NEXT: ret float %add
+
+; ORDER2: define float @caller(float %arg) {
+; ORDER2-NEXT: %result = call float @declared_as_nonan(float %arg)
+; ORDER2-NEXT: ret float %result
+
+
+declare nofpclass(nan) float @declared_as_nonan(float nofpclass(nan))
+
+define float @caller(float %arg) {
+  %result = call float @declared_as_nonan(float %arg)
+  ret float %result
+}