Allow CuDNN 7.2 to convert float -> fp16 for tensor core ops

[ThisIsIsaac]

Please make sure that this is a feature request. As per our GitHub Policy, we only address code/doc bugs, performance issues, feature requests and build/installation issues on GitHub. tag:feature_template

System information

• TensorFlow version (you are using): Tensorflow using cuDNN 7.2 and above
• Are you willing to contribute it (Yes/No): Yes

Describe the feature and the current behavior/state.
Now, if users want, cuDNN can convert fp32 inputs to fp16 to allow using tensor cores even when inputs are fp32. Before, users had to explicitly convert fp32 to fp16 using AMP. Link. This is as simple as setting CUDNN_TENSOR_OP_MATH_ALLOW_CONVERSION enum value and passing it to cudnnSetConvolutionMathType() functioin.

cudnnSetConvolutionMathType(cudnnConvDesc, CUDNN_TENSOR_OP_MATH_ALLOW_CONVERSION)

and this is the counterpart for RNN:

cudnnSetRNNMatrixMathType(cudnnRnnDesc, CUDNN_TENSOR_OP_MATH_ALLOW_CONVERSION)

We should let users chose if they want this option, since it involves converting fp32 to fp16.

[ThisIsIsaac]

ping!

This would allow 5 ~ 10x speedup for most convolutions.

@chsigg What do you think?

[sanjoy]

CC @reedwm

We (TensorFlow team) don't have cycles to actually implement this ourselves, but please feel free to create a PR.

[reedwm]

This probably would have been a good idea, but I don't think it's worth doing anymore since TensorFloat-32 is superior. TF32 has the advantage in that there's no risk of underflow or overflow compared to fp32, and gradients tend to underflow on the backwards pass in fp16 unless loss scaling is used. Therefore I'm closing this issue.

[robcaulk]

Any chance we can reopen this request considering TF32 is only supported by Ampere, where as the CUDNN_TENSOR_OP_MATH_ALLOW_CONVERSION is supported by Volta and Turing?

[ThisIsIsaac]

If it can help the performance on non-Ampere GPUs, it seems worth awhile. Would love to contribute!