Port some floating point options to new option marshalling infrastructure

[jansvoboda11]

Summary

This changes cc1 semantics for some options such as -cl-fast-relaxed-math that implied other options. Now the driver always emits all the implied options, and each option maps to one key path in CompilerInvocation so that the new option parsing system can be used.

[Anastasia]

clang/include/clang/Driver/Options.td

1220

could this also be OptInFFlag?

clang/lib/Driver/ToolChains/Clang.cpp

2830

Is this a bug fix ?

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

2

Not clear why do you need to pass these extra flags now?

[dang]

clang/include/clang/Driver/Options.td

1220

The aim was to keep the driver semantics the same as before and this was not something you could control with the driver, so I left it as just a CC1 flag. However if it makes sense to be able to control this from the driver then we can definitely make this OptInFFLag.

clang/lib/Driver/ToolChains/Clang.cpp

2830

No, in current trunk approximating floating point functions was something that was implied by other optimization flags, i.e. disabling math errno, enabling associative/reciprocal math, disabling signed zeros and disabling trapping math and -ffast-math which does all the previously mentioned things. This patch moves this logic in the driver by introducing a new CC1 flag for this so that parsing CC1 options can be more easily automated. This just reflects the logic that was previously inside cc1.

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

2

Previously passing -ffast-math to CC1 implied all these other flags. I am trying to make CC1 option parsing as simple as possible, so that we can then make it easy to generate a command line from a CompilerInvocation instance. You can refer to http://lists.llvm.org/pipermail/cfe-dev/2020-May/065421.html for more details on why we want to be able to do this

[Anastasia]

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

2

Just to understand, there used to be implied flags and it made the manual command line use of clang more compact and easy... Is the idea now to change those compound flags such that individul flags always need to be passed?

Although I thought you are still adding the implicit flags:

{options::OPT_cl_fast_relaxed_math,
 [&](const Arg *Arg) {
   RenderArg(Arg);

   CmdArgs.push_back(GetArgString(options::OPT_cl_mad_enable));
   CmdArgs.push_back(GetArgString(options::OPT_ffast_math));
   CmdArgs.push_back(GetArgString(options::OPT_ffinite_math_only));
   CmdArgs.push_back(
       GetArgString(options::OPT_menable_unsafe_fp_math));
   CmdArgs.push_back(GetArgString(options::OPT_mreassociate));
   CmdArgs.push_back(GetArgString(options::OPT_menable_no_nans));
   CmdArgs.push_back(
       GetArgString(options::OPT_menable_no_infinities));
   CmdArgs.push_back(GetArgString(options::OPT_fno_signed_zeros));
   CmdArgs.push_back(GetArgString(options::OPT_freciprocal_math));
   CmdArgs.push_back(GetArgString(options::OPT_fapprox_func));
 }}

Do I just misunderstand something?

[dang]

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

2

The command line of the driver doesn't change. This patch only affects what CC1 understands, now CC1 doesn't know anymore that -cl-fast-relaxed-math implies all these other options so the driver is responsible for specifying them when it constructs the CC1 command line.

To summarize, the clang driver command line isn't affected by this patch and it shouldn't be so let me know if something is wrong there. However, manually constructed clang -cc1 invocations need to specify the all the implied flags manually now.

[Anastasia]

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

2

Yes I understand, however, I am wondering whether this is intuitive because it seems the behavior of clang with -cc1 and without will be different if the same -cl-fast-relaxed-math flag is passed.

I also find adding all the flags manually is too verbode if -cl-fast-relaxed-math assumes to enable all the extra setting.

[dang]

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

2

My understanding is that -cc1 is an internal interface, so end-users should never use -cc1 directly and/or rely on itss interface. It is already the case that flags mean very different things to the driver and -cc1 for example "--target=" and "-triple". Furthermore, this impacted very few tests which leads me to believe that few compiler developers actually rely on this behavior.

Do you think this would be a major inconvenience to compiler developers to have to manually expand it out?

[dang]

Ping?

[jansvoboda11]

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

2

Hi @Anastasia, I'll be taking over this patch. I agree with Daniel that -cc1 is an internal interface that doesn't need to match the public driver interface.
The current approach is by far the simplest to get command-line option marshaling working.

What are your thoughts?

[jansvoboda11]

Rebase onto master.

[Anastasia]

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

2

Sorry for the delay.

My understanding is that -cc1 is an internal interface, so end-users should never use -cc1 directly and/or rely on itss interface.

This is true in practice but there are developers that use -cc1 too. My main concern is however that the internal testing gets more complicated - with so many more flags to be added that can also be easy to miss.

Is there any compromise we could find?

[Anastasia]

Just to clarify aside from the concern I have raised regarding internal testing I am not in any strong opposition of this feature. So if the community decides that it is more important to have this feature than to keep the tests short I am fine with this. However, due to the limited time I will not be able to continue reviewing this change as I don't want to block the progress due to my priorities.

[dexonsmith]

I have an idea: use DEFAULT_VALUE to keep current behaviour. Here's an example to demonstrate.

cl_mad_enable is implied by OPT_cl_unsafe_math_optimizations or OPT_cl_fast_relaxed_math. Instead of setting the default value to "false" it could be set to "CodeGenOpts.CLUnsafeMath || LangOpts.FastRelaxedMath" (one hitch is that CLUnsafeMath doesn't currently exist).

You'd need to update the OPTION_WITH_MARSHALLING_FLAG definition in parseSimpleArgs to something like:

#define OPTION_WITH_MARSHALLING_FLAG(PREFIX_TYPE, NAME, ID, KIND, GROUP,       \
                                     ALIAS, ALIASARGS, FLAGS, PARAM, HELPTEXT, \
                                     METAVAR, VALUES, SPELLING, ALWAYS_EMIT,   \
                                     KEYPATH, DEFAULT_VALUE, IS_POSITIVE)      \
  this->KEYPATH = (Args.hasArg(OPT_##ID) && IS_POSITIVE) || (DEFAULT_VALUE);

@Bigcheese / @jansvoboda11 , WDYT?

Edit: Note that generating the -cc1 should already mostly work, you'd just want to add () around DEFAULT_VALUE:

#define OPTION_WITH_MARSHALLING_FLAG(PREFIX_TYPE, NAME, ID, KIND, GROUP,       \
                                     ALIAS, ALIASARGS, FLAGS, PARAM, HELPTEXT, \
                                     METAVAR, VALUES, SPELLING, ALWAYS_EMIT,   \
                                     KEYPATH, DEFAULT_VALUE, IS_POSITIVE)      \
  if ((FLAGS) & options::CC1Option &&                                            \
      (ALWAYS_EMIT || this->KEYPATH != (DEFAULT_VALUE)))                         \
    Args.push_back(SPELLING);

[dexonsmith]

In D82756#2348159, @dexonsmith wrote:

I have an idea: use DEFAULT_VALUE to keep current behaviour. Here's an example to demonstrate.

Downside of this is that it's error prone, because it requires for correctness (but does not enforce) a particular option definition order. The alternative would be to formalize it in the options somehow.

[jansvoboda11]

@dexonsmith, thanks for sharing your idea. Overall, I like the simplicity.

As you say, not every command line option has an associated CodeGenOptions or LangOptions field. For this patch only, we'd need to create the following fields:

• CLUnsafeOptimizations for OPT_cl_unsafe_math_optimizations
• CLFiniteMathOnly for OPT_cl_finite_math_only
• CLNoSignedZeros for OPT_cl_no_signed_zeros

I'm not sure how many other cases there are outside of what this patch touches.

Introducing ordering-sensitivity is not ideal, but if all options are defined in close proximity to each other, it should be relatively easy to reason about.

Correct me if I'm wrong, but when generating the command line, all "implied" flags would be hidden, even if they were explicit in the original comand line:

• original command line: clang -cc1 -cl-unsafe-math-optimizations -cl-mad-enable -menable-unsafe-fp-math -mreassociate -fno-signed-zeros -freciprocal-math -fapprox-func [...]
• generated command line: clang -cc1 -cl-unsafe-math-optimizations [...]

This might be a bit surprising, but I don't think this would cause issues for explicit modules. What are your thoughts?

Formalizing the "implies" relationships would make it possible to remove the ordering-sensitivity and possibly generate implied flags even when explicitly passed to cc1. It would complicate the TableGen backend, which I'd prefer to keep as simple as possible.

[dexonsmith]

In D82756#2350233, @jansvoboda11 wrote:

Correct me if I'm wrong, but when generating the command line, all "implied" flags would be hidden, even if they were explicit in the original comand line:

• original command line: clang -cc1 -cl-unsafe-math-optimizations -cl-mad-enable -menable-unsafe-fp-math -mreassociate -fno-signed-zeros -freciprocal-math -fapprox-func [...]
• generated command line: clang -cc1 -cl-unsafe-math-optimizations [...]

This might be a bit surprising, but I don't think this would cause issues for explicit modules. What are your thoughts?

I think this is fine. It's similar to a case where the caller might explicitly specify the default value, and it'll get canonicalized out.

Formalizing the "implies" relationships would make it possible to remove the ordering-sensitivity and possibly generate implied flags even when explicitly passed to cc1. It would complicate the TableGen backend, which I'd prefer to keep as simple as possible.

I was't thinking of dropping the ordering-sensitivity. Instead, you could just error if the referenced option hadn't declared already. One idea would be to change the tablegen to something like:

MarshallingInfoFlag<
    "CodeGenOpts.LessPreciseFPMAD",
    DefaultAnyOf<[cl_unsafe_math_optimizations, cl_fast_relaxed_math]>>;

in the definition of cl_mad_enable, then:

• error if they aren't defined first; and
• construct a default value out of the key-paths.

I think this would less error-prone for maintenance, since it designs away some really subtle bugs.

[dexonsmith]

IOW, the goal of formalizing would just be to:

• error if the .td file defined options in the wrong order to get correct parsing
• automatically generate the code for default value, instead of having to re-type the name of the keypath

[jansvoboda11]

Formalize the "implied by" relationship with DefaultAnyOf<[...]>.

[jansvoboda11]

Using the records with DefaultAnyOf<[...]> to catch errors seems to work well, thanks for the suggestion.

(I was initially thinking about using the keypaths to avoid errors and reorder the options in the TableGen backend to make everything work.)

@dexonsmith WDYT about the implementation?

@Anastasia There is no need to manually expand -cc1 options anymore. Do you have any other concerns?

[jansvoboda11]

Fix typo & whitespace.

[dexonsmith]

I like how this is coming together. I have a few comments inline.

Also, I wonder if there should be a test for the new OptParser behaviour in llvm/unittests/Option/.

clang/include/clang/Driver/Options.td

1220

I think adding a driver flag (if that's the right thing to do) should be done separately in a follow-up commit.

Also for a separate commit: it would be a great improvement if you could have OptIn / OptOut flags that were -cc1-only (maybe CC1OptInFFlag).

• Both -fX and -fno-X would be parsed by -cc1 (and cancel each other out).
• Only the non-default one would be generated when serializing to -cc1 from CompilerInvocation (for OptIn, we'd never generate -fno-X).
• Neither would be recognized by the driver.

I suggest we might want that for most -cc11 flags. This would make it easier to poke through the driver with -Xclang to override -cc1 options the driver adds. Not something we want for end-users, but useful for debugging the compiler itself. Currently the workflow is to run the driver with -###, copy/paste, search for and remove the option you want to skip, and finally run the -cc1 command...

The reason I bring it up is that it's possible people will start using OptInFFLag just in order to get this behaviour, not because they intend to add a driver flag.

clang/lib/Frontend/CompilerInvocation.cpp

3706

I don't have an opinion about whether there should be a newline here, but please make unrelated whitespace changes like this in a separate commit (before/after).

llvm/utils/TableGen/OptParserEmitter.cpp

456–460

I think array_pod_sort will use this like a bool, similar to std::sort, in which case you I think you want:

return (*A)->getID() < (*B)->getID();

464–465

I'm curious if this is necessary. If so, how do the options get out-of-order?

Also, a cleaner way to call array_pod_sort would be:

llvm::sort(OptsWithMarshalling, CmpMarshallingOpts);

and I would be tempted to define the lambda inline in the call to llvm::sort.

If it's not necessary, I suggest replacing with an assertion:

assert(llvm::is_sorted(OptsWithMarshalling, ...));

[jansvoboda11]

Thanks for the feedback Duncan.

I don't think this patch introduces any changes the parser. We only change the way how CodeGenOpts and LangOpts get populated when using DefaultAnyOf<[...]>. I've added a test of CompilerInvocation that checks just that.

clang/include/clang/Driver/Options.td

1220

I agree that making all OptInFFlag and OptOutFFlag driver flags as well as -cc1 flags by default is not great. How would we go about deciding which options are okay to demote to -cc1-only? Perhaps those not present in ClangCommandLineReference.rst and driver invocations in tests?

clang/lib/Frontend/CompilerInvocation.cpp

3706

Got it.

llvm/utils/TableGen/OptParserEmitter.cpp

456–460

I see that array_pod_sort calls qsort from the C standard library, which should use the result of comparator as an int.

464–465

1. The options get out of order during parsing. The RecordKeeper stores records in std::map<std::string, std::unique_ptr<Record>, std::less<>> that maintains lexicographical order.

2. Later, they are reordered in this function before prefix groups are generated: array_pod_sort(Opts.begin(), Opts.end(), CompareOptionRecords);.

3. Before we generate the marshalling code, we need to restore the definition order so that we don't use a LangOpts or CodeGenOpts field (from DefaultAnyOf) before it was initialized.

I've added more detailed explanation to the comment.

I used array_pod_sort to be consistent with what's already used here in OptParserEmitter.cpp. I will switch to llvm::sort to be more concise if we don't mind the potential code bloat described here https://llvm.org/doxygen/namespacellvm.html#add1eb5637dd671428b6f138ed3db6428.

[jansvoboda11]

Integrated code review suggestions.

[dexonsmith]

In D82756#2361565, @jansvoboda11 wrote:

Thanks for the feedback Duncan.

I don't think this patch introduces any changes the parser. We only change the way how CodeGenOpts and LangOpts get populated when using DefaultAnyOf<[...]>. I've added a test of CompilerInvocation that checks just that.

The test for CompilerInvocation looks great, but IMO it's insufficient.

Given that the changes are in llvm/, it seems best to have a test there so that check-llvm (also) catches any breakage. I took a look at llvm/unittests/Option/Opts.td and llvm/unittests/Option/OptionParsingTest.cpp and I see we don't currently have any tests for marshalling, but my intuition is it wouldn't be hard to do. What I suggest is adding OptionMarshallingTest.cpp and just test the new behaviour from this commit (key properties of the changes you made to OptParser.td and OptParserEmitter.cpp), leaving testing the rest for some follow-up.

clang/include/clang/Driver/Options.td

1220

How would we go about deciding which options are okay to demote to -cc1-only?

The key is not to add (or remove) driver options unintentionally. Driver options are clang's public interface, and once an option shows up there we're supposed to support it "forever". We shouldn't be accidentally/incidentally growing that surface area in order to simplify parsing/generating -cc1 command-lines.

I based my comment on @dang's reason for not using OptInFFLag, which I agree with:

The aim was to keep the driver semantics the same as before and this was not something you could control with the driver, so I left it as just a CC1 flag.

llvm/utils/TableGen/OptParserEmitter.cpp

456–460

Thanks, you're right, I misremembered array_pod_sort somehow reinterpreting the lambda...

464–465

Thanks for the explanation about the ordering, this makes sense.

Regarding array_pod_sort, I was referring to how llvm::sort calls array_pod_sort when it can... but I hadn't noticed before that it can't do this for a custom comparator. You should stick with array_pod_sort (although maybe as a follow-up I'll look into whether we can detect if the custom comparator to llvm::sort is stateless and defer to array_pod_sort in that case as well...)

[jansvoboda11]

Added LLVM unit tests, reverted back to array_pod_sort.

[jansvoboda11]

I've added tests for OptParserEmitter. Let me know if you had something more detailed in mind.

clang/include/clang/Driver/Options.td

1220

Agreed.

llvm/utils/TableGen/OptParserEmitter.cpp

464–465

No problem. I've switched back to array_pod_sort.

[dexonsmith]

LGTM, thanks for working through this.

[jansvoboda11]

Thanks for the feedback.

@Bigcheese do you have anything to add?

[Bigcheese]

LGTM

[jansvoboda11]

@dexonsmith, could you please commit this one for me? I don't have the rights to do so, as this is my first patch.

[dexonsmith]

In D82756#2382885, @jansvoboda11 wrote:

@dexonsmith, could you please commit this one for me? I don't have the rights to do so, as this is my first patch.

Sure; what do you want for GIT_AUTHOR_NAME and GIT_AUTHOR_EMAIL?

[jansvoboda11]

Thanks, Jan Svoboda and [e-mail] is fine. Is it possible to add @dang as a co-author? git log says he uses Daniel Grumberg and [e-mail].

[dexonsmith]

In D82756#2383317, @jansvoboda11 wrote:

Thanks, Jan Svoboda and [e-mail] is fine. Is it possible to add @dang as a co-author? git log says he uses Daniel Grumberg and [e-mail].

I don't think there's a Git feature to support that, but I'll credit Daniel in the commit message:

Port some floating point options to new option marshalling infrastructure

This ports a number of OpenCL and fast-math flags for floating point
over to the new marshalling infrastructure.

As part of this, Opt{In,Out}FFlag were enhanced to allow other flags to
imply them, via DefaultAnyOf<>. For example:

defm signed_zeros : OptOutFFlag<"signed-zeros", ...,
  "LangOpts->NoSignedZero",
  DefaultAnyOf<[cl_no_signed_zeros, menable_unsafe_fp_math]>>;

defines -fsigned-zeros (false) and -fno-signed-zeros (true)
linked to the keypath LangOpts->NoSignedZero, defaulting to false,
but set to true implicitly if one of -cl-no-signed-zeros or
-menable-unsafe-fp-math is on.

Note that the initial patch was written Daniel Grumberg.

I just rebased and I'm building and running tests now... should get back to check on it and commit later today at some point.

[jansvoboda11]

Thank you!

clang/include/clang/Driver/Options.td

 // Group that ignores all gcc optimizations that won't be implemented
 def clang_ignored_gcc_optimization_f_Group : OptionGroup<
   "<clang_ignored_gcc_optimization_f_Group>">, Group<f_Group>, Flags<[Ignored]>;
 
 // A boolean option which is opt-in in CC1. The positive option exists in CC1 and
 // Args.hasArg(OPT_ffoo) is used to check that the flag is enabled.
 // This is useful if the option is usually disabled.
 multiclass OptInFFlag<string name, string pos_prefix, string neg_prefix="",
-                      string help="", list<OptionFlag> flags=[]> {
+                      string help="", list<OptionFlag> flags=[], code keypath=""> {
   def f#NAME : Flag<["-"], "f"#name>, Flags<!listconcat([CC1Option], flags)>,
-               Group<f_Group>, HelpText<!strconcat(pos_prefix, help)>;
+               Group<f_Group>, HelpText<!strconcat(pos_prefix, help)>,
+               MarshallingInfoFlag<keypath, "false">;
   def fno_#NAME : Flag<["-"], "fno-"#name>, Flags<flags>,
                Group<f_Group>, HelpText<!strconcat(neg_prefix, help)>;
 }
 
 // A boolean option which is opt-out in CC1. The negative option exists in CC1 and
 // Args.hasArg(OPT_fno_foo) is used to check that the flag is disabled.
 multiclass OptOutFFlag<string name, string pos_prefix, string neg_prefix,
-                       string help="", list<OptionFlag> flags=[]> {
+                       string help="", list<OptionFlag> flags=[], code keypath=""> {
   def f#NAME : Flag<["-"], "f"#name>, Flags<flags>,
                Group<f_Group>, HelpText<!strconcat(pos_prefix, help)>;
   def fno_#NAME : Flag<["-"], "fno-"#name>, Flags<!listconcat([CC1Option], flags)>,
-               Group<f_Group>, HelpText<!strconcat(neg_prefix, help)>;
+               Group<f_Group>, HelpText<!strconcat(neg_prefix, help)>,
+               MarshallingInfoFlag<keypath, "false">;
 }
 
 /////////
 // Options
 
 // The internal option ID must be a valid C++ identifier and results in a
 // clang::driver::options::OPT_XX enum constant for XX.
 //

 def bundle__loader : Separate<["-"], "bundle_loader">;
 def bundle : Flag<["-"], "bundle">;
 def b : JoinedOrSeparate<["-"], "b">, Flags<[Unsupported]>;
 def cl_opt_disable : Flag<["-"], "cl-opt-disable">, Group<opencl_Group>, Flags<[CC1Option]>,
   HelpText<"OpenCL only. This option disables all optimizations. By default optimizations are enabled.">;
 def cl_strict_aliasing : Flag<["-"], "cl-strict-aliasing">, Group<opencl_Group>, Flags<[CC1Option]>,
   HelpText<"OpenCL only. This option is added for compatibility with OpenCL 1.0.">;
 def cl_single_precision_constant : Flag<["-"], "cl-single-precision-constant">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Treat double precision floating-point constant as single precision constant.">;
-def cl_finite_math_only : Flag<["-"], "cl-finite-math-only">, Group<opencl_Group>, Flags<[CC1Option]>,
+  HelpText<"OpenCL only. Treat double precision floating-point constant as single precision constant.">,
+  MarshallingInfoFlag<"LangOpts->SinglePrecisionConstants", "false">;
+def cl_finite_math_only : Flag<["-"], "cl-finite-math-only">, Group<opencl_Group>,
   HelpText<"OpenCL only. Allow floating-point optimizations that assume arguments and results are not NaNs or +-Inf.">;
 def cl_kernel_arg_info : Flag<["-"], "cl-kernel-arg-info">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Generate kernel argument metadata.">;
-def cl_unsafe_math_optimizations : Flag<["-"], "cl-unsafe-math-optimizations">, Group<opencl_Group>, Flags<[CC1Option]>,
+  HelpText<"OpenCL only. Generate kernel argument metadata.">,
+  MarshallingInfoFlag<"CodeGenOpts.EmitOpenCLArgMetadata", "false">;
+def cl_unsafe_math_optimizations : Flag<["-"], "cl-unsafe-math-optimizations">, Group<opencl_Group>,
   HelpText<"OpenCL only. Allow unsafe floating-point optimizations.  Also implies -cl-no-signed-zeros and -cl-mad-enable.">;
 def cl_fast_relaxed_math : Flag<["-"], "cl-fast-relaxed-math">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Sets -cl-finite-math-only and -cl-unsafe-math-optimizations, and defines __FAST_RELAXED_MATH__.">;
+  HelpText<"OpenCL only. Sets -cl-finite-math-only and -cl-unsafe-math-optimizations, and defines __FAST_RELAXED_MATH__.">,
+  MarshallingInfoFlag<"LangOpts->FastRelaxedMath", "false">;
 def cl_mad_enable : Flag<["-"], "cl-mad-enable">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Allow use of less precise MAD computations in the generated binary.">;
-def cl_no_signed_zeros : Flag<["-"], "cl-no-signed-zeros">, Group<opencl_Group>, Flags<[CC1Option]>,
+  HelpText<"OpenCL only. Allow use of less precise MAD computations in the generated binary.">,
+  MarshallingInfoFlag<"CodeGenOpts.LessPreciseFPMAD", "false">;
+def cl_no_signed_zeros : Flag<["-"], "cl-no-signed-zeros">, Group<opencl_Group>,
   HelpText<"OpenCL only. Allow use of less precise no signed zeros computations in the generated binary.">;
 def cl_std_EQ : Joined<["-"], "cl-std=">, Group<opencl_Group>, Flags<[CC1Option]>,
   HelpText<"OpenCL language standard to compile for.">, Values<"cl,CL,cl1.1,CL1.1,cl1.2,CL1.2,cl2.0,CL2.0,cl3.0,CL3.0,clc++,CLC++">;
 def cl_denorms_are_zero : Flag<["-"], "cl-denorms-are-zero">, Group<opencl_Group>,
   HelpText<"OpenCL only. Allow denormals to be flushed to zero.">;
 def cl_fp32_correctly_rounded_divide_sqrt : Flag<["-"], "cl-fp32-correctly-rounded-divide-sqrt">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Specify that single precision floating-point divide and sqrt used in the program source are correctly rounded.">;
+  HelpText<"OpenCL only. Specify that single precision floating-point divide and sqrt used in the program source are correctly rounded.">,
+  MarshallingInfoFlag<"CodeGenOpts.CorrectlyRoundedDivSqrt", "false">;
 def cl_uniform_work_group_size : Flag<["-"], "cl-uniform-work-group-size">, Group<opencl_Group>, Flags<[CC1Option]>,
-  HelpText<"OpenCL only. Defines that the global work-size be a multiple of the work-group size specified to clEnqueueNDRangeKernel">;
+  HelpText<"OpenCL only. Defines that the global work-size be a multiple of the work-group size specified to clEnqueueNDRangeKernel">,
+  MarshallingInfoFlag<"CodeGenOpts.UniformWGSize", "false">;
 def client__name : JoinedOrSeparate<["-"], "client_name">;
 def combine : Flag<["-", "--"], "combine">, Flags<[DriverOption, Unsupported]>;
 def compatibility__version : JoinedOrSeparate<["-"], "compatibility_version">;
 def config : Separate<["--"], "config">, Flags<[DriverOption]>,
   HelpText<"Specifies configuration file">;
 def config_system_dir_EQ : Joined<["--"], "config-system-dir=">, Flags<[DriverOption, HelpHidden]>,
   HelpText<"System directory for configuration files">;
 def config_user_dir_EQ : Joined<["--"], "config-user-dir=">, Flags<[DriverOption, HelpHidden]>,

 def : Flag<["-"], "fextended-identifiers">, Group<clang_ignored_f_Group>;
 def : Flag<["-"], "fno-extended-identifiers">, Group<f_Group>, Flags<[Unsupported]>;
 def fhosted : Flag<["-"], "fhosted">, Group<f_Group>;
 def fdenormal_fp_math_EQ : Joined<["-"], "fdenormal-fp-math=">, Group<f_Group>, Flags<[CC1Option]>;
 def ffp_model_EQ : Joined<["-"], "ffp-model=">, Group<f_Group>, Flags<[DriverOption]>,
   HelpText<"Controls the semantics of floating-point calculations.">;
 def ffp_exception_behavior_EQ : Joined<["-"], "ffp-exception-behavior=">, Group<f_Group>, Flags<[CC1Option]>,
   HelpText<"Specifies the exception behavior of floating-point operations.">;
-defm fast_math : OptInFFlag<"fast-math", "Allow aggressive, lossy floating-point optimizations">;
+defm fast_math : OptInFFlag<"fast-math", "Allow aggressive, lossy floating-point, optimizations", "", "", [], "LangOpts->FastMath">;
 defm math_errno : OptInFFlag<"math-errno", "Require math functions to indicate errors by setting errno">;
 def fbracket_depth_EQ : Joined<["-"], "fbracket-depth=">, Group<f_Group>, Flags<[CoreOption]>;
 def fsignaling_math : Flag<["-"], "fsignaling-math">, Group<f_Group>;
 def fno_signaling_math : Flag<["-"], "fno-signaling-math">, Group<f_Group>;
 defm jump_tables : OptOutFFlag<"jump-tables", "Use", "Do not use", " jump tables for lowering switches">;
 defm force_enable_int128 : OptInFFlag<"force-enable-int128", "Enable", "Disable", " support for int128_t type">;
 defm keep_static_consts : OptInFFlag<"keep-static-consts", "Keep", "Don't keep", " static const variables if unused", [DriverOption]>;
 defm fixed_point : OptInFFlag<"fixed-point", "Enable", "Disable", " fixed point types">;

 } // end -f[no-]sanitize* flags
 
 def funsafe_math_optimizations : Flag<["-"], "funsafe-math-optimizations">,
   Group<f_Group>;
 def fno_unsafe_math_optimizations : Flag<["-"], "fno-unsafe-math-optimizations">,
   Group<f_Group>;
 def fassociative_math : Flag<["-"], "fassociative-math">, Group<f_Group>;
 def fno_associative_math : Flag<["-"], "fno-associative-math">, Group<f_Group>;
-def freciprocal_math :
-  Flag<["-"], "freciprocal-math">, Group<f_Group>, Flags<[CC1Option]>,
-  HelpText<"Allow division operations to be reassociated">;
-def fno_reciprocal_math : Flag<["-"], "fno-reciprocal-math">, Group<f_Group>;
-def ffinite_math_only : Flag<["-"], "ffinite-math-only">, Group<f_Group>, Flags<[CC1Option]>;
-def fno_finite_math_only : Flag<["-"], "fno-finite-math-only">, Group<f_Group>;
-def fsigned_zeros : Flag<["-"], "fsigned-zeros">, Group<f_Group>;
-def fno_signed_zeros :
-  Flag<["-"], "fno-signed-zeros">, Group<f_Group>, Flags<[CC1Option]>,
-  HelpText<"Allow optimizations that ignore the sign of floating point zeros">;
+defm reciprocal_math : OptInFFlag< "reciprocal-math", "Allow division operations to be reassociated", "", "", [], "LangOpts->AllowRecip">;
+def fapprox_func : Flag<["-"], "fapprox-func">, Group<f_Group>, Flags<[CC1Option, NoDriverOption]>,
+  MarshallingInfoFlag<"LangOpts->ApproxFunc", "false">;
+defm finite_math_only : OptInFFlag<"finite-math-only", "", "", "", [], "LangOpts->FiniteMathOnly">;
+defm signed_zeros : OptOutFFlag<"signed-zeros", "Allow optimizations that ignore the sign of floating point zeros", "", "", [], "LangOpts->NoSignedZero">;
 def fhonor_nans : Flag<["-"], "fhonor-nans">, Group<f_Group>;
 def fno_honor_nans : Flag<["-"], "fno-honor-nans">, Group<f_Group>;
 def fhonor_infinities : Flag<["-"], "fhonor-infinities">, Group<f_Group>;
 def fno_honor_infinities : Flag<["-"], "fno-honor-infinities">, Group<f_Group>;
 // This option was originally misspelt "infinites" [sic].
 def : Flag<["-"], "fhonor-infinites">, Alias<fhonor_infinities>;
 def : Flag<["-"], "fno-honor-infinites">, Alias<fno_honor_infinities>;
 def frounding_math : Flag<["-"], "frounding-math">, Group<f_Group>, Flags<[CC1Option]>;
def new_struct_path_tbaa : Flag<["-"], "new-struct-path-tbaa">,
   HelpText<"Enable enhanced struct-path aware Type Based Alias Analysis">;
 def mdebug_pass : Separate<["-"], "mdebug-pass">,
   HelpText<"Enable additional debug output">;
 def mframe_pointer_EQ : Joined<["-"], "mframe-pointer=">,
   HelpText<"Specify which frame pointers to retain (all, non-leaf, none).">, Values<"all,non-leaf,none">;
 def mdisable_tail_calls : Flag<["-"], "mdisable-tail-calls">,
   HelpText<"Disable tail call optimization, keeping the call stack accurate">;
 def menable_no_infinities : Flag<["-"], "menable-no-infs">,
-  HelpText<"Allow optimization to assume there are no infinities.">;
+  HelpText<"Allow optimization to assume there are no infinities.">,
+  MarshallingInfoFlag<"LangOpts->NoHonorInfs", "false">;
 def menable_no_nans : Flag<["-"], "menable-no-nans">,
-  HelpText<"Allow optimization to assume there are no NaNs.">;
+  HelpText<"Allow optimization to assume there are no NaNs.">,
+  MarshallingInfoFlag<"LangOpts->NoHonorNaNs", "false">;
 def menable_unsafe_fp_math : Flag<["-"], "menable-unsafe-fp-math">,
   HelpText<"Allow unsafe floating-point math optimizations which may decrease "
-           "precision">;
+           "precision">,
+  MarshallingInfoFlag<"LangOpts->UnsafeFPMath", "false">;
 def mreassociate : Flag<["-"], "mreassociate">,
-  HelpText<"Allow reassociation transformations for floating-point instructions">;
+  HelpText<"Allow reassociation transformations for floating-point instructions">,
+  MarshallingInfoFlag<"LangOpts->AllowFPReassoc", "false">;
 def mabi_EQ_ieeelongdouble : Flag<["-"], "mabi=ieeelongdouble">,
   HelpText<"Use IEEE 754 quadruple-precision for long double">;
 def mfloat_abi : Separate<["-"], "mfloat-abi">,
   HelpText<"The float ABI to use">;
 def mtp : Separate<["-"], "mtp">,
   HelpText<"Mode for reading thread pointer">;
 def mlimit_float_precision : Separate<["-"], "mlimit-float-precision">,
   HelpText<"Limit float precision to the given value">;

clang/lib/Driver/ToolChains/Clang.cpp

const llvm::DenormalMode DefaultDenormalFPMath =
       TC.getDefaultDenormalModeForType(Args, JA);
   const llvm::DenormalMode DefaultDenormalFP32Math =
       TC.getDefaultDenormalModeForType(Args, JA, &llvm::APFloat::IEEEsingle());
 
   llvm::DenormalMode DenormalFPMath = DefaultDenormalFPMath;
   llvm::DenormalMode DenormalFP32Math = DefaultDenormalFP32Math;
   StringRef FPContract = "";
   bool StrictFPModel = false;
-
+  bool ApproxFunc = false;
 
   if (const Arg *A = Args.getLastArg(options::OPT_flimited_precision_EQ)) {
     CmdArgs.push_back("-mlimit-float-precision");
     CmdArgs.push_back(A->getValue());
   }
 
   for (const Arg *A : Args) {
     auto optID = A->getOption().getID();
case options::OPT_ffast_math:
       MathErrno = false;
       AssociativeMath = true;
       ReciprocalMath = true;
       SignedZeros = false;
       TrappingMath = false;
       RoundingFPMath = false;
       // If fast-math is set then set the fp-contract mode to fast.
       FPContract = "fast";
+      ApproxFunc = true;
       break;
     case options::OPT_fno_fast_math:
       HonorINFs = true;
       HonorNaNs = true;
       // Turning on -ffast-math (with either flag) removes the need for
       // MathErrno. However, turning *off* -ffast-math merely restores the
       // toolchain default (which may be false).
       MathErrno = TC.IsMathErrnoDefault();
static void RenderFloatingPointOptions(const ToolChain &TC, const Driver &D,
 
   if (!HonorNaNs)
     CmdArgs.push_back("-menable-no-nans");
 
   if (MathErrno)
     CmdArgs.push_back("-fmath-errno");
 
   if (!MathErrno && AssociativeMath && ReciprocalMath && !SignedZeros &&
-      !TrappingMath)
+      !TrappingMath) {
     CmdArgs.push_back("-menable-unsafe-fp-math");
+    ApproxFunc = true;
+  }
+
+  if (ApproxFunc)
+    CmdArgs.push_back("-fapprox-func");
 
   if (!SignedZeros)
     CmdArgs.push_back("-fno-signed-zeros");
 
   if (AssociativeMath && !SignedZeros && !TrappingMath)
     CmdArgs.push_back("-mreassociate");
 
   if (ReciprocalMath)
if (Arg *A =
     if (std::stoi(Val.str()) <= 0)
       D.Diag(diag::err_drv_trivial_auto_var_init_stop_after_invalid_value);
     CmdArgs.push_back(
         Args.MakeArgString("-ftrivial-auto-var-init-stop-after=" + Val));
   }
 }
 
 static void RenderOpenCLOptions(const ArgList &Args, ArgStringList &CmdArgs) {
+  const auto &OptTable = getDriverOptTable();
+
+  auto GetArgString = [&Args, &OptTable](OptSpecifier Specifier) {
+    return Args.MakeArgString(OptTable.getOption(Specifier).getPrefixedName());
+  };
+
+  auto RenderArg = [&](const Arg *Arg) {
+    CmdArgs.push_back(Args.MakeArgString(Arg->getOption().getPrefixedName()));
+  };
+
+  auto RenderFiniteMathOnlyArg = [&](const Arg *Arg) {
+    CmdArgs.push_back(GetArgString(options::OPT_ffinite_math_only));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_no_nans));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_no_infinities));
+  };
+
+  auto RenderUnsafeMathOptsArg = [&](const Arg *Arg) {
+    CmdArgs.push_back(GetArgString(options::OPT_cl_mad_enable));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_unsafe_fp_math));
+    CmdArgs.push_back(GetArgString(options::OPT_mreassociate));
+    CmdArgs.push_back(GetArgString(options::OPT_fno_signed_zeros));
+    CmdArgs.push_back(GetArgString(options::OPT_freciprocal_math));
+    CmdArgs.push_back(GetArgString(options::OPT_fapprox_func));
+  };
+
+  auto RenderFastRelaxedMathArg = [&](const Arg *Arg) {
+    RenderArg(Arg);
+
+    CmdArgs.push_back(GetArgString(options::OPT_cl_mad_enable));
+    CmdArgs.push_back(GetArgString(options::OPT_ffast_math));
+    CmdArgs.push_back(GetArgString(options::OPT_ffinite_math_only));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_unsafe_fp_math));
+    CmdArgs.push_back(GetArgString(options::OPT_mreassociate));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_no_nans));
+    CmdArgs.push_back(GetArgString(options::OPT_menable_no_infinities));
+    CmdArgs.push_back(GetArgString(options::OPT_fno_signed_zeros));
+    CmdArgs.push_back(GetArgString(options::OPT_freciprocal_math));
+    CmdArgs.push_back(GetArgString(options::OPT_fapprox_func));
+  };
+
+  auto RenderNoSignedZerosArg = [&](const Arg *Arg) {
+    CmdArgs.push_back(GetArgString(options::OPT_fno_signed_zeros));
+  };
+
   // cl-denorms-are-zero is not forwarded. It is translated into a generic flag
   // for denormal flushing handling based on the target.
-  const unsigned ForwardedArguments[] = {
-      options::OPT_cl_opt_disable,
-      options::OPT_cl_strict_aliasing,
-      options::OPT_cl_single_precision_constant,
-      options::OPT_cl_finite_math_only,
-      options::OPT_cl_kernel_arg_info,
-      options::OPT_cl_unsafe_math_optimizations,
-      options::OPT_cl_fast_relaxed_math,
-      options::OPT_cl_mad_enable,
-      options::OPT_cl_no_signed_zeros,
-      options::OPT_cl_fp32_correctly_rounded_divide_sqrt,
-      options::OPT_cl_uniform_work_group_size
-  };
+  const std::pair<unsigned, llvm::function_ref<void(const Arg *Arg)>>
+      CLArguments[] = {
+          {options::OPT_cl_opt_disable, RenderArg},
+          {options::OPT_cl_strict_aliasing, RenderArg},
+          {options::OPT_cl_single_precision_constant, RenderArg},
+          {options::OPT_cl_finite_math_only, RenderFiniteMathOnlyArg},
+          {options::OPT_cl_kernel_arg_info, RenderArg},
+          {options::OPT_cl_unsafe_math_optimizations, RenderUnsafeMathOptsArg},
+          {options::OPT_cl_fast_relaxed_math, RenderFastRelaxedMathArg},
+          {options::OPT_cl_mad_enable, RenderArg},
+          {options::OPT_cl_no_signed_zeros, RenderNoSignedZerosArg},
+          {options::OPT_cl_fp32_correctly_rounded_divide_sqrt, RenderArg},
+          {options::OPT_cl_uniform_work_group_size, RenderArg}};
 
   if (Arg *A = Args.getLastArg(options::OPT_cl_std_EQ)) {
     std::string CLStdStr = std::string("-cl-std=") + A->getValue();
     CmdArgs.push_back(Args.MakeArgString(CLStdStr));
   }
 
-  for (const auto &Arg : ForwardedArguments)
-    if (const auto *A = Args.getLastArg(Arg))
-      CmdArgs.push_back(Args.MakeArgString(A->getOption().getPrefixedName()));
+  for (const auto &Arg : CLArguments)
+    if (const auto *A = Args.getLastArg(Arg.first))
+      Arg.second(A);
 }
 
 static void RenderARCMigrateToolOptions(const Driver &D, const ArgList &Args,
                                         ArgStringList &CmdArgs) {
   bool ARCMTEnabled = false;
   if (!Args.hasArg(options::OPT_fno_objc_arc, options::OPT_fobjc_arc)) {
     if (const Arg *A = Args.getLastArg(options::OPT_ccc_arcmt_check,
                                        options::OPT_ccc_arcmt_modify,

clang/lib/Frontend/CompilerInvocation.cpp

static bool ParseCodeGenArgs(CodeGenOptions &Opts, ArgList &Args, InputKind IK,
   }
 
   Opts.DisableFree = Args.hasArg(OPT_disable_free);
   Opts.DiscardValueNames = Args.hasArg(OPT_discard_value_names);
   Opts.DisableTailCalls = Args.hasArg(OPT_mdisable_tail_calls);
   Opts.NoEscapingBlockTailCalls =
       Args.hasArg(OPT_fno_escaping_block_tail_calls);
   Opts.FloatABI = std::string(Args.getLastArgValue(OPT_mfloat_abi));
-  Opts.LessPreciseFPMAD = Args.hasArg(OPT_cl_mad_enable) ||
-                          Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                          Args.hasArg(OPT_cl_fast_relaxed_math);
   Opts.LimitFloatPrecision =
       std::string(Args.getLastArgValue(OPT_mlimit_float_precision));
-  Opts.CorrectlyRoundedDivSqrt =
-      Args.hasArg(OPT_cl_fp32_correctly_rounded_divide_sqrt);
-  Opts.UniformWGSize =
-      Args.hasArg(OPT_cl_uniform_work_group_size);
   Opts.Reciprocals = Args.getAllArgValues(OPT_mrecip_EQ);
   Opts.StrictFloatCastOverflow =
       !Args.hasArg(OPT_fno_strict_float_cast_overflow);
 
   Opts.NoZeroInitializedInBSS = Args.hasArg(OPT_fno_zero_initialized_in_bss);
   Opts.NumRegisterParameters = getLastArgIntValue(Args, OPT_mregparm, 0, Diags);
   Opts.NoExecStack = Args.hasArg(OPT_mno_exec_stack);
   Opts.SmallDataLimit =
#include "clang/Basic/LangStandards.def"
   Opts.PIE = Args.hasArg(OPT_pic_is_pie);
   Opts.Static = Args.hasArg(OPT_static_define);
   Opts.DumpRecordLayoutsSimple = Args.hasArg(OPT_fdump_record_layouts_simple);
   Opts.DumpRecordLayouts = Opts.DumpRecordLayoutsSimple
                         || Args.hasArg(OPT_fdump_record_layouts);
   Opts.DumpVTableLayouts = Args.hasArg(OPT_fdump_vtable_layouts);
   Opts.SpellChecking = !Args.hasArg(OPT_fno_spell_checking);
   Opts.NoBitFieldTypeAlign = Args.hasArg(OPT_fno_bitfield_type_align);
-  Opts.SinglePrecisionConstants = Args.hasArg(OPT_cl_single_precision_constant);
-  Opts.FastRelaxedMath = Args.hasArg(OPT_cl_fast_relaxed_math);
   if (Opts.FastRelaxedMath)
     Opts.setDefaultFPContractMode(LangOptions::FPM_Fast);
   Opts.HexagonQdsp6Compat = Args.hasArg(OPT_mqdsp6_compat);
   Opts.FakeAddressSpaceMap = Args.hasArg(OPT_ffake_address_space_map);
   Opts.ParseUnknownAnytype = Args.hasArg(OPT_funknown_anytype);
   Opts.DebuggerSupport = Args.hasArg(OPT_fdebugger_support);
   Opts.DebuggerCastResultToId = Args.hasArg(OPT_fdebugger_cast_result_to_id);
   Opts.DebuggerObjCLiteral = Args.hasArg(OPT_fdebugger_objc_literal);
#include "clang/Basic/LangStandards.def"
   // inlining enabled.
   Opts.NoInlineDefine = !Opts.Optimize;
   if (Arg *InlineArg = Args.getLastArg(
           options::OPT_finline_functions, options::OPT_finline_hint_functions,
           options::OPT_fno_inline_functions, options::OPT_fno_inline))
     if (InlineArg->getOption().matches(options::OPT_fno_inline))
       Opts.NoInlineDefine = true;
 
-  Opts.FastMath =
-      Args.hasArg(OPT_ffast_math) || Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.FiniteMathOnly = Args.hasArg(OPT_ffinite_math_only) ||
-                        Args.hasArg(OPT_ffast_math) ||
-                        Args.hasArg(OPT_cl_finite_math_only) ||
-                        Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.UnsafeFPMath = Args.hasArg(OPT_menable_unsafe_fp_math) ||
-                      Args.hasArg(OPT_ffast_math) ||
-                      Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                      Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.AllowFPReassoc = Args.hasArg(OPT_mreassociate) ||
-                        Args.hasArg(OPT_menable_unsafe_fp_math) ||
-                        Args.hasArg(OPT_ffast_math) ||
-                        Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                        Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.NoHonorNaNs =
-      Args.hasArg(OPT_menable_no_nans) || Args.hasArg(OPT_ffinite_math_only) ||
-      Args.hasArg(OPT_ffast_math) || Args.hasArg(OPT_cl_finite_math_only) ||
-      Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.NoHonorInfs = Args.hasArg(OPT_menable_no_infinities) ||
-                     Args.hasArg(OPT_ffinite_math_only) ||
-                     Args.hasArg(OPT_ffast_math) ||
-                     Args.hasArg(OPT_cl_finite_math_only) ||
-                     Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.NoSignedZero = Args.hasArg(OPT_fno_signed_zeros) ||
-                      Args.hasArg(OPT_menable_unsafe_fp_math) ||
-                      Args.hasArg(OPT_ffast_math) ||
-                      Args.hasArg(OPT_cl_no_signed_zeros) ||
-                      Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                      Args.hasArg(OPT_cl_fast_relaxed_math);
-  Opts.AllowRecip = Args.hasArg(OPT_freciprocal_math) ||
-                    Args.hasArg(OPT_menable_unsafe_fp_math) ||
-                    Args.hasArg(OPT_ffast_math) ||
-                    Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                    Args.hasArg(OPT_cl_fast_relaxed_math);
-  // Currently there's no clang option to enable this individually
-  Opts.ApproxFunc = Args.hasArg(OPT_menable_unsafe_fp_math) ||
-                    Args.hasArg(OPT_ffast_math) ||
-                    Args.hasArg(OPT_cl_unsafe_math_optimizations) ||
-                    Args.hasArg(OPT_cl_fast_relaxed_math);
-
   if (Arg *A = Args.getLastArg(OPT_ffp_contract)) {
     StringRef Val = A->getValue();
     if (Val == "fast")
       Opts.setDefaultFPContractMode(LangOptions::FPM_Fast);
     else if (Val == "on")
       Opts.setDefaultFPContractMode(LangOptions::FPM_On);
     else if (Val == "off")
       Opts.setDefaultFPContractMode(LangOptions::FPM_Off);
if (auto MaybeValue = NORMALIZER(OPT_##ID, TABLE_INDEX, Args, Diags))      \
       this->KEYPATH = static_cast<TYPE>(*MaybeValue);                          \
     else                                                                       \
       this->KEYPATH = DEFAULT_VALUE;                                           \
   }
 
 #include "clang/Driver/Options.inc"
 #undef OPTION_WITH_MARSHALLING_STRING
 #undef OPTION_WITH_MARSHALLING_FLAG
   return true;
 }
 
 bool CompilerInvocation::CreateFromArgs(CompilerInvocation &Res,
                                         ArrayRef<const char *> CommandLineArgs,
                                         DiagnosticsEngine &Diags,
                                         const char *Argv0) {
   bool Success = true;
 

clang/test/CodeGen/complex-math.c

 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple x86_64-unknown-unknown -o - | FileCheck %s --check-prefix=X86
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple x86_64-pc-win64 -o - | FileCheck %s --check-prefix=X86
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple i686-unknown-unknown -o - | FileCheck %s --check-prefix=X86
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple powerpc-unknown-unknown -o - | FileCheck %s --check-prefix=PPC
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple armv7-none-linux-gnueabi -o - | FileCheck %s --check-prefix=ARM
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple armv7-none-linux-gnueabihf -o - | FileCheck %s --check-prefix=ARMHF
 // RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple thumbv7k-apple-watchos2.0 -o - -target-abi aapcs16 | FileCheck %s --check-prefix=ARM7K
-// RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple aarch64-unknown-unknown -ffast-math -ffp-contract=fast -o - | FileCheck %s --check-prefix=AARCH64-FASTMATH
+// RUN: %clang_cc1 %s -O0 -fno-experimental-new-pass-manager -emit-llvm -triple aarch64-unknown-unknown -ffast-math \
+// RUN:    -menable-unsafe-fp-math -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math -fapprox-func \
+// RUN:    -mreassociate -ffinite-math-only -ffp-contract=fast -o - | FileCheck %s --check-prefix=AARCH64-FASTMATH
 
 float _Complex add_float_rr(float a, float b) {
   // X86-LABEL: @add_float_rr(
   // X86: fadd
   // X86-NOT: fadd
   // X86: ret
   return a + b;
 }

clang/test/CodeGen/fast-math.c

-// RUN: %clang_cc1 -ffast-math -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s
+// RUN: %clang_cc1 -ffast-math -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math \
+// RUN:   -menable-unsafe-fp-math -ffinite-math-only -fapprox-func -mreassociate -ffp-contract=fast \
+// RUN:   -emit-llvm -o - %s | FileCheck %s
+
 float f0, f1, f2;
 
 void foo(void) {
   // CHECK-LABEL: define {{.*}}void @foo()
 
   // CHECK: fadd fast
   f0 = f1 + f2;
 
   // CHECK: ret
 }

clang/test/CodeGen/finite-math.c

-// RUN: %clang_cc1 -ffinite-math-only -emit-llvm -o - %s | FileCheck %s -check-prefix=CHECK -check-prefix=FINITE
+// RUN: %clang_cc1 -ffinite-math-only -menable-no-nans -menable-no-infs -emit-llvm -o - \
+// RUN:   %s | FileCheck %s -check-prefix=CHECK -check-prefix=FINITE
 // RUN: %clang_cc1 -fno-signed-zeros -emit-llvm -o - %s | FileCheck %s -check-prefix=CHECK  -check-prefix=NSZ
 // RUN: %clang_cc1 -freciprocal-math -emit-llvm -o - %s | FileCheck %s -check-prefix=CHECK  -check-prefix=RECIP
 // RUN: %clang_cc1 -mreassociate -emit-llvm -o - %s | FileCheck %s -check-prefix=CHECK  -check-prefix=REASSOC
 
 float f0, f1, f2;
 
 void foo(void) {
   // CHECK-LABEL: define {{.*}}void @foo()

clang/test/CodeGen/fp-floatcontrol-stack.cpp

 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffp-contract=on -DDEFAULT=1 -emit-llvm -o - %s | FileCheck --check-prefix=CHECK-DDEFAULT %s
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffp-contract=on -DEBSTRICT=1 -ffp-exception-behavior=strict -emit-llvm -o - %s | FileCheck --check-prefix=CHECK-DEBSTRICT %s
-// RUN: %clang_cc1 -triple x86_64-linux-gnu -DFAST=1 -ffast-math -ffp-contract=fast -emit-llvm -o - %s | FileCheck --check-prefix=CHECK-FAST %s
+// RUN: %clang_cc1 -triple x86_64-linux-gnu -DFAST=1 -ffast-math -ffinite-math-only -menable-unsafe-fp-math \
+// RUN:   -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math -fapprox-func -mreassociate \
+// RUN:   -ffp-contract=fast -emit-llvm -o - %s | FileCheck --check-prefix=CHECK-FAST %s
 // RUN: %clang_cc1 -triple x86_64-linux-gnu -ffp-contract=on -DNOHONOR=1 -menable-no-infs -menable-no-nans -emit-llvm -o - %s | FileCheck --check-prefix=CHECK-NOHONOR %s
 
 #define FUN(n) \
   (float z) { return n * z + n; }
 
 float fun_default FUN(1)
 //CHECK-LABEL: define {{.*}} @_Z11fun_defaultf{{.*}}
 #if DEFAULT

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 -ffinite-math-only -menable-unsafe-fp-math \
+// RUN:   -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math \
+// RUN:   -fapprox-func -mreassociate -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s
 
 float test_default(float a, float b, float c) {
   float tmp = a;
   tmp += b;
   tmp += c;
   return tmp;
 }
 

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

 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -emit-llvm -o - %s | FileCheck -check-prefix CHECK-PRECISE %s
 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -menable-no-nans -emit-llvm -o - %s | FileCheck -check-prefix CHECK-NO-NANS %s
 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -menable-no-infs -emit-llvm -o - %s | FileCheck -check-prefix CHECK-NO-INFS %s
-// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ffinite-math-only -emit-llvm -o - %s | FileCheck -check-prefix CHECK-FINITE %s
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ffinite-math-only -menable-no-infs -menable-no-nans -emit-llvm -o - %s | FileCheck -check-prefix CHECK-FINITE %s
 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -fno-signed-zeros -emit-llvm -o - %s | FileCheck -check-prefix CHECK-NO-SIGNED-ZEROS %s
 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -mreassociate -emit-llvm -o - %s | FileCheck -check-prefix CHECK-REASSOC %s
 // RUN: %clang_cc1 -triple x86_64-unknown-unknown -freciprocal-math -emit-llvm -o - %s | FileCheck -check-prefix CHECK-RECIP %s
-// RUN: %clang_cc1 -triple x86_64-unknown-unknown -menable-unsafe-fp-math -emit-llvm -o - %s | FileCheck -check-prefix CHECK-UNSAFE %s
-// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ffast-math -emit-llvm -o - %s | FileCheck -check-prefix CHECK-FAST %s
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -menable-unsafe-fp-math -freciprocal-math \
+// RUN:   -mreassociate -fno-signed-zeros -fapprox-func -emit-llvm -o - %s | FileCheck -check-prefix CHECK-UNSAFE %s
+// RUN: %clang_cc1 -triple x86_64-unknown-unknown -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans \
+// RUN:   -fno-signed-zeros -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -emit-llvm \
+// RUN:   -o - %s | FileCheck -check-prefix CHECK-FAST %s
 
 float fn(float);
 
 float test(float a) {
   return a + fn(a);
 }
 
 // CHECK-PRECISE: [[CALL_RES:%.+]] = call float @fn(float {{%.+}})

clang/test/CodeGen/fpconstrained.c

 // RUN: %clang_cc1 -ftrapping-math -frounding-math -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=FPMODELSTRICT
 // RUN: %clang_cc1 -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s -check-prefix=PRECISE
-// RUN: %clang_cc1 -ffast-math -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 -ffast-math -emit-llvm -o - %s | FileCheck %s -check-prefix=FASTNOCONTRACT
-// RUN: %clang_cc1 -ffast-math -ffp-contract=fast -ffp-exception-behavior=ignore -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 -ffast-math -ffp-contract=fast -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=EXCEPT
-// RUN: %clang_cc1 -ffast-math -ffp-contract=fast -ffp-exception-behavior=maytrap -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=MAYTRAP
-
+// RUN: %clang_cc1 -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast \
+// RUN:   -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate \
+// RUN:   -emit-llvm -o - %s | FileCheck %s -check-prefix=FASTNOCONTRACT
+// RUN: %clang_cc1 -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast \
+// RUN:   -ffp-exception-behavior=ignore -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast \
+// RUN:   -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm -o - %s \
+// RUN:   | FileCheck %s -check-prefix=EXCEPT
+// RUN: %clang_cc1 -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast \
+// RUN:   -ffp-exception-behavior=maytrap -fexperimental-strict-floating-point -emit-llvm -o - %s \
+// RUN:   | FileCheck %s -check-prefix=MAYTRAP
+
 float f0, f1, f2;
 
 void foo() {
   // CHECK-LABEL: define {{.*}}void @foo()
 
   // MAYTRAP: llvm.experimental.constrained.fadd.f32(float %{{.*}}, float %{{.*}}, metadata !"round.tonearest", metadata !"fpexcept.maytrap")
   // EXCEPT: llvm.experimental.constrained.fadd.f32(float %{{.*}}, float %{{.*}}, metadata !"round.tonearest", metadata !"fpexcept.strict")
   // FPMODELSTRICT: llvm.experimental.constrained.fadd.f32(float %{{.*}}, float %{{.*}}, metadata !"round.dynamic", metadata !"fpexcept.strict")

clang/test/CodeGen/fpconstrained.cpp

 // RUN: %clang_cc1 -x c++ -ftrapping-math -fexceptions -fcxx-exceptions -frounding-math -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=FPMODELSTRICT
 // RUN: %clang_cc1 -x c++ -ffp-contract=fast -fexceptions -fcxx-exceptions -emit-llvm -o - %s | FileCheck %s -check-prefix=PRECISE
-// RUN: %clang_cc1 -x c++ -ffast-math -fexceptions -fcxx-exceptions -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 -x c++ -ffast-math -fexceptions -fcxx-exceptions -emit-llvm -o - %s | FileCheck %s -check-prefix=FASTNOCONTRACT
-// RUN: %clang_cc1 -x c++ -ffast-math -fexceptions -fcxx-exceptions -ffp-contract=fast -ffp-exception-behavior=ignore -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 -x c++ -ffast-math -fexceptions -fcxx-exceptions -ffp-contract=fast -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=EXCEPT
-// RUN: %clang_cc1 -x c++ -ffast-math -fexceptions -fcxx-exceptions -ffp-contract=fast -ffp-exception-behavior=maytrap -fexperimental-strict-floating-point -emit-llvm -o - %s | FileCheck %s -check-prefix=MAYTRAP
-
+// RUN: %clang_cc1 -x c++ -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -fexceptions -fcxx-exceptions \
+// RUN:   -ffp-contract=fast -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 -x c++ -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -fexceptions -fcxx-exceptions \
+// RUN:   -emit-llvm -o - %s | FileCheck %s -check-prefix=FASTNOCONTRACT
+// RUN: %clang_cc1 -x c++ -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -fexceptions -fcxx-exceptions \
+// RUN:   -ffp-contract=fast -ffp-exception-behavior=ignore -emit-llvm -o - %s | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 -x c++ -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -fexceptions -fcxx-exceptions \
+// RUN:   -ffp-contract=fast -ffp-exception-behavior=strict -fexperimental-strict-floating-point -emit-llvm \
+// RUN:   -o - %s | FileCheck %s -check-prefix=EXCEPT
+// RUN: %clang_cc1 -x c++ -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros \
+// RUN:   -freciprocal-math -menable-unsafe-fp-math -fapprox-func -mreassociate -fexceptions -fcxx-exceptions \
+// RUN:   -ffp-contract=fast -ffp-exception-behavior=maytrap -fexperimental-strict-floating-point -emit-llvm \
+// RUN:   -o - %s | FileCheck %s -check-prefix=MAYTRAP
+
 float f0, f1, f2;
 
   template <class>
   class aaaa {
    public:
     ~aaaa();
     void b();
   };

clang/test/CodeGen/libcalls.c

 // RUN: %clang_cc1 -fmath-errno -emit-llvm -o - %s -triple i386-unknown-unknown | FileCheck -check-prefix CHECK-YES %s
 // RUN: %clang_cc1 -emit-llvm -o - %s -triple i386-unknown-unknown | FileCheck -check-prefix CHECK-NO %s
-// RUN: %clang_cc1 -menable-unsafe-fp-math -emit-llvm -o - %s -triple i386-unknown-unknown | FileCheck -check-prefix CHECK-FAST %s
+// RUN: %clang_cc1 -menable-unsafe-fp-math -fapprox-func -fno-signed-zeros -mreassociate -freciprocal-math \
+// RUN:   -emit-llvm -o - %s -triple i386-unknown-unknown | FileCheck -check-prefix CHECK-FAST %s
 
 // CHECK-YES-LABEL: define void @test_sqrt
 // CHECK-NO-LABEL: define void @test_sqrt
 // CHECK-FAST-LABEL: define void @test_sqrt
 void test_sqrt(float a0, double a1, long double a2) {
   // CHECK-YES: call float @sqrtf
   // CHECK-NO: call float @llvm.sqrt.f32(float
   // CHECK-FAST: call reassoc nsz arcp afn float @llvm.sqrt.f32(float

clang/test/CodeGenOpenCL/no-signed-zeros.cl

 // RUN: %clang_cc1 %s -emit-llvm -o - | FileCheck %s -check-prefix=NORMAL
-// RUN: %clang_cc1 %s -emit-llvm -cl-no-signed-zeros -o - | FileCheck %s -check-prefix=NO-SIGNED-ZEROS
+// RUN: %clang_cc1 %s -emit-llvm -fno-signed-zeros -o - | FileCheck %s -check-prefix=NO-SIGNED-ZEROS
 
 float signedzeros(float a) {
   return a;
 }
 
 // CHECK: attributes
 // NORMAL: "no-signed-zeros-fp-math"="false"
 // NO-SIGNED-ZEROS: "no-signed-zeros-fp-math"="true"

clang/test/CodeGenOpenCL/relaxed-fpmath.cl

 // RUN: %clang_cc1 %s -emit-llvm -o - | FileCheck %s -check-prefix=NORMAL
-// RUN: %clang_cc1 %s -emit-llvm -cl-fast-relaxed-math -o - | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 %s -emit-llvm -cl-finite-math-only -o - | FileCheck %s -check-prefix=FINITE
-// RUN: %clang_cc1 %s -emit-llvm -cl-unsafe-math-optimizations -o - | FileCheck %s -check-prefix=UNSAFE
+// RUN: %clang_cc1 %s -emit-llvm -cl-fast-relaxed-math -cl-mad-enable -ffast-math -ffinite-math-only \
+// RUN:   -menable-unsafe-fp-math -mreassociate -menable-no-nans -menable-no-infs -fno-signed-zeros \
+// RUN:   -freciprocal-math -fapprox-func -o - | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 %s -emit-llvm -ffinite-math-only -menable-no-nans -menable-no-infs -o - | FileCheck %s -check-prefix=FINITE
+// RUN: %clang_cc1 %s -emit-llvm -cl-mad-enable -menable-unsafe-fp-math -mreassociate -fno-signed-zeros \
+// RUN:   -freciprocal-math -fapprox-func -o - | FileCheck %s -check-prefix=UNSAFE
 // RUN: %clang_cc1 %s -emit-llvm -cl-mad-enable -o - | FileCheck %s -check-prefix=MAD
-// RUN: %clang_cc1 %s -emit-llvm -cl-no-signed-zeros -o - | FileCheck %s -check-prefix=NOSIGNED
+// RUN: %clang_cc1 %s -emit-llvm -fno-signed-zeros -o - | FileCheck %s -check-prefix=NOSIGNED
 
 // Check the fp options are correct with PCH.
 // RUN: %clang_cc1 %s -DGEN_PCH=1 -finclude-default-header -triple spir-unknown-unknown -emit-pch -o %t.pch
 // RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -o - | FileCheck %s -check-prefix=NORMAL
-// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-fast-relaxed-math -o - | FileCheck %s -check-prefix=FAST
-// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-finite-math-only -o - | FileCheck %s -check-prefix=FINITE
-// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-unsafe-math-optimizations -o - | FileCheck %s -check-prefix=UNSAFE
+// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-fast-relaxed-math -cl-mad-enable -ffast-math \
+// RUN:   -ffinite-math-only -menable-unsafe-fp-math -mreassociate -menable-no-nans -menable-no-infs -fno-signed-zeros \
+// RUN:   -freciprocal-math -fapprox-func -o - | FileCheck %s -check-prefix=FAST
+// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -ffinite-math-only -menable-no-nans -menable-no-infs -o - \
+// RUN:   | FileCheck %s -check-prefix=FINITE
+// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-mad-enable -menable-unsafe-fp-math -mreassociate \
+// RUN:   -fno-signed-zeros -freciprocal-math -fapprox-func -o - | FileCheck %s -check-prefix=UNSAFE
 // RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-mad-enable -o - | FileCheck %s -check-prefix=MAD
-// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -cl-no-signed-zeros -o - | FileCheck %s -check-prefix=NOSIGNED
+// RUN: %clang_cc1 %s -include-pch %t.pch -fno-validate-pch -emit-llvm -fno-signed-zeros -o - | FileCheck %s -check-prefix=NOSIGNED
 
 #if !GEN_PCH
 
 float spscalardiv(float a, float b) {
   // CHECK: @spscalardiv(
 
   // NORMAL: fdiv float
   // FAST: fdiv fast float

clang/test/Driver/opencl.cl

 // CHECK-CL11: "-cc1" {{.*}} "-cl-std=CL1.1"
 // CHECK-CL12: "-cc1" {{.*}} "-cl-std=CL1.2"
 // CHECK-CL20: "-cc1" {{.*}} "-cl-std=CL2.0"
 // CHECK-CL30: "-cc1" {{.*}} "-cl-std=CL3.0"
 // CHECK-CLCPP: "-cc1" {{.*}} "-cl-std=clc++"
 // CHECK-OPT-DISABLE: "-cc1" {{.*}} "-cl-opt-disable"
 // CHECK-STRICT-ALIASING: "-cc1" {{.*}} "-cl-strict-aliasing"
 // CHECK-SINGLE-PRECISION-CONST: "-cc1" {{.*}} "-cl-single-precision-constant"
-// CHECK-FINITE-MATH-ONLY: "-cc1" {{.*}} "-cl-finite-math-only"
+// CHECK-FINITE-MATH-ONLY: "-cc1" {{.*}} "-ffinite-math-only" "-menable-no-nans" "-menable-no-infs"
 // CHECK-KERNEL-ARG-INFO: "-cc1" {{.*}} "-cl-kernel-arg-info"
-// CHECK-UNSAFE-MATH-OPT: "-cc1" {{.*}} "-cl-unsafe-math-optimizations"
-// CHECK-FAST-RELAXED-MATH: "-cc1" {{.*}} "-cl-fast-relaxed-math"
+// CHECK-UNSAFE-MATH-OPT: "-cc1" {{.*}} "-cl-mad-enable" "-menable-unsafe-fp-math" "-mreassociate" "-fno-signed-zeros" "-freciprocal-math" "-fapprox-func"
+// CHECK-FAST-RELAXED-MATH: "-cc1" {{.*}} "-cl-fast-relaxed-math" "-cl-mad-enable" "-ffast-math" "-ffinite-math-only" "-menable-unsafe-fp-math" "-mreassociate" "-menable-no-nans" "-menable-no-infs" "-fno-signed-zeros" "-freciprocal-math" "-fapprox-func"
 // CHECK-MAD-ENABLE: "-cc1" {{.*}} "-cl-mad-enable"
-// CHECK-NO-SIGNED-ZEROS: "-cc1" {{.*}} "-cl-no-signed-zeros"
+// CHECK-NO-SIGNED-ZEROS: "-cc1" {{.*}} "-fno-signed-zeros"
 
 // This is not forwarded
 // CHECK-DENORMS-ARE-ZERO-NOT: "-cl-denorms-are-zero"
 
 // CHECK-ROUND-DIV: "-cc1" {{.*}} "-cl-fp32-correctly-rounded-divide-sqrt"
 // CHECK-UNIFORM-WG: "-cc1" {{.*}} "-cl-uniform-work-group-size"
 // CHECK-C99: error: invalid value 'c99' in '-cl-std=c99'
 // CHECK-INVALID: error: invalid value 'invalid' in '-cl-std=invalid'
 
 kernel void func(void);

clang/test/Headers/nvptx_device_math_sin.c

 // REQUIRES: nvptx-registered-target
 // RUN: %clang_cc1 -x c -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
 // RUN: %clang_cc1 -x c -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --check-prefix=SLOW
-// RUN: %clang_cc1 -x c -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -ffast-math -ffp-contract=fast
-// RUN: %clang_cc1 -x c -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - -ffast-math -ffp-contract=fast | FileCheck %s --check-prefix=FAST
+// RUN: %clang_cc1 -x c -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown \
+// RUN:   -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -ffast-math -ffinite-math-only \
+// RUN:   -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math -menable-unsafe-fp-math \
+// RUN:   -fapprox-func -mreassociate -ffp-contract=fast
+// RUN: %clang_cc1 -x c -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers \
+// RUN:   -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown \
+// RUN:   -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc \
+// RUN:   -o - -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math \
+// RUN:   -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast | FileCheck %s --check-prefix=FAST
 // expected-no-diagnostics
 
 #include <math.h>
 
 double math(float f, double d) {
   double r = 0;
 // SLOW:  call float @__nv_sinf(float
 // FAST:  call fast float @__nv_fast_sinf(float

clang/test/Headers/nvptx_device_math_sin.cpp

 // REQUIRES: nvptx-registered-target
 // RUN: %clang_cc1 -x c++ -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
 // RUN: %clang_cc1 -x c++ -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --check-prefix=SLOW
-// RUN: %clang_cc1 -x c++ -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -ffast-math -ffp-contract=fast
-// RUN: %clang_cc1 -x c++ -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - -ffast-math -ffp-contract=fast | FileCheck %s --check-prefix=FAST
+// RUN: %clang_cc1 -x c++ -internal-isystem %S/Inputs/include -fopenmp -triple powerpc64le-unknown-unknown \
+// RUN:   -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -ffast-math -ffinite-math-only \
+// RUN:   -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math -menable-unsafe-fp-math \
+// RUN:   -fapprox-func -mreassociate -ffp-contract=fast
+// RUN: %clang_cc1 -x c++ -include __clang_openmp_device_functions.h -internal-isystem %S/../../lib/Headers/openmp_wrappers \
+// RUN:   -internal-isystem %S/Inputs/include -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown \
+// RUN:   -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc \
+// RUN:   -o - -ffast-math -ffinite-math-only -menable-no-infs -menable-no-nans -fno-signed-zeros -freciprocal-math \
+// RUN:   -menable-unsafe-fp-math -fapprox-func -mreassociate -ffp-contract=fast | FileCheck %s --check-prefix=FAST
 // expected-no-diagnostics
 
 #include <cmath>
 
 double math(float f, double d) {
   double r = 0;
 // SLOW:  call float @__nv_sinf(float
 // FAST:  call fast float @__nv_fast_sinf(float

clang/test/SemaOpenCL/fp-options.cl

 // RUN: %clang_cc1 %s -finclude-default-header -triple spir-unknown-unknown -emit-pch -o %t.pch
-// RUN: %clang_cc1 %s -finclude-default-header -cl-no-signed-zeros -triple spir-unknown-unknown -include-pch %t.pch -fsyntax-only -verify
+// RUN: %clang_cc1 %s -finclude-default-header -fno-signed-zeros -triple spir-unknown-unknown -include-pch %t.pch -fsyntax-only -verify
 // expected-no-diagnostics
 

llvm/utils/TableGen/OptParserEmitter.cpp

void EmitOptParser(RecordKeeper &Records, raw_ostream &OS) {
   std::vector<std::unique_ptr<MarshallingKindInfo>> OptsWithMarshalling;
   for (unsigned I = 0, E = Opts.size(); I != E; ++I) {
     const Record &R = *Opts[I];
 
     // Start a single option entry.
     OS << "OPTION(";
     WriteOptRecordFields(OS, R);
     OS << ")\n";
-    if (!isa<UnsetInit>(R.getValueInit("MarshallingKind")))
+    if (!isa<UnsetInit>(R.getValueInit("MarshallingKind")) &&
+        !R.getValueAsString("KeyPath").empty())
       OptsWithMarshalling.push_back(MarshallingKindInfo::create(R));
   }
   OS << "#endif // OPTION\n";
 
   for (const auto &KindInfo : OptsWithMarshalling) {
     OS << "#ifdef " << KindInfo->MacroName << "\n";
     OS << KindInfo->MacroName << "(";
     WriteOptRecordFields(OS, KindInfo->R);
     OS << ", ";
     KindInfo->emit(OS);
     OS << ")\n";
     OS << "#endif // " << KindInfo->MacroName << "\n";
   }
 
   OS << "\n";
   OS << "#ifdef SIMPLE_ENUM_VALUE_TABLE";
   OS << "\n";
   OS << MarshallingStringInfo::ValueTablePreamble;
   std::vector<StringRef> ValueTableNames;
   for (const auto &KindInfo : OptsWithMarshalling)
     if (auto MaybeValueTableName = KindInfo->emitValueTable(OS))
       ValueTableNames.push_back(*MaybeValueTableName);
 
   OS << MarshallingStringInfo::ValueTablesDecl << "{";
   for (auto ValueTableName : ValueTableNames)
     OS << "{" << ValueTableName << ", sizeof(" << ValueTableName