[Build] Build fails: 'error : no operator "+=" matches these operands' with nv_bfloat16

[AndreyOrb]

Describe the issue

Build fails on the latest version: SHA-1: 3a47bd2

No operator "+=" matches operands
nv_bfloat16 += nv_bfloat16
nv_bfloat16 += const nv_bfloat16
nv_bfloat16 += float

Urgency

Urgent! Build with CUDA 11.8 fails

Target platform

Windows

Build script

.\build.bat --config Debug --build_shared_lib --parallel --use_cuda --cuda_home "C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.8" --cudnn_home "C:\Program Files\NVIDIA GPU Computing Toolkit\cuDNN\cudnn-windows-x86_64-8.9.0.131_cuda11-archive" --cuda_version 11.8 --use_tensorrt --tensorrt_home "C:\Program Files\NVIDIA GPU Computing Toolkit\TensorRT\TensorRT-10.9.0.34.Windows.win10.cuda-11.8\TensorRT-10.9.0.34" --use_tensorrt_oss_parser --cmake_generator "Visual Studio 16 2019" --compile_no_warning_as_error --cmake_path E:\3rdParties\cmake-4.0.3\build\bin\Release\cmake.exe --skip_tests --enable_cuda_line_info --use_mimalloc

Workaround of building without contrib ops (--disable_contrib_ops) does not work in my case:

CMake Error at onnxruntime_providers_tensorrt.cmake:4 (message):
  To compile TensorRT execution provider contrib ops have to be enabled to
  dump an engine using com.microsoft:EPContext node.
Call Stack (most recent call first):
  onnxruntime_providers.cmake:132 (include)
  CMakeLists.txt:1890 (include)

Error / output

E:\3rdParties\onnxruntime_v1.22.0\onnxruntime\contrib_ops\cuda\bert\skip_layer_norm_impl.cu(167): error : no operator "+=" matches these operands [E:\3rdParties\onnxruntime_v1.22.0\build\Windows\Debug\onnxrun
time_providers_cuda.vcxproj]
operand types are: nv_bfloat16 += nv_bfloat16
detected during:
instantiation of "void onnxruntime::contrib::cuda::SkipLayerNormKernelSmall<T,TPB,ILP,Simplified>(T *, T *, const T *, const T *, const T *, const T *, const T *, T, int, int) [with T=nv_bfloat1
6, TPB=32U, ILP=4, Simplified=true]"
(240): here
instantiation of "void onnxruntime::contrib::cuda::LaunchSkipLayerNormKernel<T,Simplified>(cudaStream_t, T *, T *, const T *, const T *, const T *, const T *, const T *, float, int, int, int) [w
ith T=nv_bfloat16, Simplified=true]"
(272): here

Visual Studio Version

VS 16 2019

GCC / Compiler Version

No response

[AndreyOrb]

nv_bfloat16 is defined in CUDA\v11.8\include\cuda_bf16.h but it is different from CUDA\v12.6\include\cuda_bf16.h
Can it be that the build fails only for CUDA 11.x, but works fine with CUDA 12.x ?

[AndreyOrb]

I think it is related to #24772
cc yuslepukhin, I can see that you approved the PR.

Do you have any tests for CUDA 11.8? It's still supported, right?
I can't check the windows test results to verify because they were not saved: https://github.com/microsoft/onnxruntime/actions/runs/15122036726/job/42506419762?pr=24772

[AndreyOrb]

I have reviewed the pipeline file. The tests run only with CUDA 12.2: https://github.com/microsoft/onnxruntime/blob/main/.github/workflows/windows_cuda.yml

[AndreyOrb]

I reverted 3a47bd2 locally and the build succeeded, so definitely there is a problem with this commit.

Is there a chance anyone can fix the nv_bfloat16 to work properly with CUDA 11.8 ?

[AndreyOrb]

@tianleiwu I see that you have been working on bfloat16 recently (#25161)
Is there anything that can be done about this issue?

[tianleiwu]

@AndreyOrb.
I'll take a look later on Monday (need time to setup environment for cuda 11.8). It shall be a simple fix from the error message.

For example, add a few lines in onnxruntime/core/providers/cuda/cu_inc/common.cuh

#if ((__CUDACC_VER_MAJOR__ == 11) && (__CUDACC_VER_MINOR__ <= 8))
__host__ __device__
__nv_bfloat16& operator+=(__nv_bfloat16 &a, const __nv_bfloat16 &b) {
    a = __hadd(a, b);
    return a;
}

__host__ __device__
__nv_bfloat16& operator+=(__nv_bfloat16 &a, const float &b) {
    a = __hadd(a, __float2bfloat16(b));
    return a;
}
#endif