编译错误：error: more than one conversion function from "half" to a built-in type applies

[Wanglongzhi2001]

• 状况： cmake 完毕后进行 ninja -j8后报错，错误信息：

  
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cu(68): error: more than one conversion function from "half" to a built-in type applies:
          function "__half::operator float() const"
  /usr/local/cuda/include/cuda_fp16.hpp(204): here
          function "__half::operator short() const"
  /usr/local/cuda/include/cuda_fp16.hpp(222): here
          function "__half::operator unsigned short() const"
  /usr/local/cuda/include/cuda_fp16.hpp(225): here
          function "__half::operator int() const"
  /usr/local/cuda/include/cuda_fp16.hpp(228): here
          function "__half::operator unsigned int() const"
  /usr/local/cuda/include/cuda_fp16.hpp(231): here
          function "__half::operator long long() const"
  /usr/local/cuda/include/cuda_fp16.hpp(234): here
          function "__half::operator unsigned long long() const"
  /usr/local/cuda/include/cuda_fp16.hpp(237): here
          function "__half::operator __nv_bool() const"
  /usr/local/cuda/include/cuda_fp16.hpp(241): here
        detected during:
          instantiation of "void rwkv::cuda::layer_norm::<unnamed>::AffineStore<SRC, DST, do_scale, do_center>::store<N>(const SRC *, int64_t, int64_t) [with SRC=float, DST=half, do_scale=true, do_center=true, N=4]" 
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cuh(645): here
          instantiation of "void rwkv::cuda::layer_norm::LayerNormBlockSMemImpl<LOAD,STORE,ComputeType,pack_size,block_size>(LOAD, STORE, int64_t, int64_t, double, ComputeType *, ComputeType *) [with LOAD=rwkv::cuda::layer_norm::DirectLoad<half, half>, STORE=rwkv::cuda::layer_norm::<unnamed>::AffineStore<float, half, true, true>, ComputeType=float, pack_size=4, block_size=128]" 
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cuh(725): here
          instantiation of "cudaError_t rwkv::cuda::layer_norm::TryDispatchLayerNormBlockSMemImplBlockSize<LOAD,STORE,ComputeType,pack_size>(cudaStream_t, LOAD, STORE, int64_t, int64_t, double, ComputeType *, ComputeType *, __nv_bool *) [with LOAD=rwkv::cuda::layer_norm::DirectLoad<half, half>, STORE=rwkv::cuda::layer_norm::<unnamed>::AffineStore<float, half, true, true>, ComputeType=float, pack_size=4]" 
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cuh(847): here
          instantiation of "cudaError_t rwkv::cuda::layer_norm::TryDispatchLayerNormBlockSMemImplPackSize<LOAD, STORE, ComputeType>::operator()(cudaStream_t, LOAD, STORE, int64_t, int64_t, double, ComputeType *, ComputeType *, __nv_bool *) [with LOAD=rwkv::cuda::layer_norm::DirectLoad<half, half>, STORE=rwkv::cuda::layer_norm::<unnamed>::AffineStore<float, half, true, true>, ComputeType=float]" 
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cuh(870): here
          instantiation of "cudaError_t rwkv::cuda::layer_norm::TryDispatchLayerNormBlockSMemImpl(cudaStream_t, LOAD, STORE, int64_t, int64_t, double, ComputeType *, ComputeType *, __nv_bool *) [with LOAD=rwkv::cuda::layer_norm::DirectLoad<half, half>, STORE=rwkv::cuda::layer_norm::<unnamed>::AffineStore<float, half, true, true>, ComputeType=float]" 
  /root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cuh(994): here
          instantiation of "std::enable_if<<expression>, cudaError_t>::type rwkv::cuda::layer_norm::DispatchLayerNorm(cudaStream_t, LOAD, STORE, int64_t, int64_t, double, ComputeType *, ComputeType *) [with LOAD=rwkv::cuda::layer_norm::DirectLoad<half, half>, STORE=rwkv::cuda::layer_norm::<unnamed>::AffineStore<float, half, true, true>, ComputeType=float]" 
  (96): here
          instantiation of "void rwkv::cuda::layer_norm::LayerNormForwardGpu(int64_t, int64_t, double, const T *, const T *, const T *, T *) [with T=half]" 
  (120): here

6 errors detected in the compilation of "/root/Dev/MyPR/faster-rwkv/kernels/cuda/layer_norm.cu". [118/286] Building CUDA object CMakeFiles/faster_rwkv_internal.dir/kernels/cuda/cat.cu.o ninja: build stopped: subcommand failed.

- OS: WSL2(Ubuntu 20.04)
- cmake version: 3.27.0
- cmake 命令: `cmake -DFR_ENABLE_CUDA=ON -DCMAKE_BUILD_TYPE=Release -GNinja ..`

[Wanglongzhi2001]

请问是我这边的问题嘛？回退到前几个 commit 的版本依旧编译报错

[AsakusaRinne]

master分支有一个被## ===================== A solution for compiling error of half caused by low cuda arch detected =======================包裹的几行代码，把那几行取消掉注释可以解决这个问题，即以下代码：

# include(FindCUDA/select_compute_arch)
# CUDA_DETECT_INSTALLED_GPUS(INSTALLED_GPU_CCS_1)
# string(STRIP "${INSTALLED_GPU_CCS_1}" INSTALLED_GPU_CCS_2)
# string(REPLACE " " "\;" INSTALLED_GPU_CCS_3 "${INSTALLED_GPU_CCS_2}")
# string(REPLACE "." "" CUDA_ARCH_LIST "${INSTALLED_GPU_CCS_3}")
# STRING(FIND "${CUDA_ARCH_LIST}" "\;" LAST_SEMICOLON_INDEX REVERSE)
# math(EXPR SPLIT_START_INDEX "${LAST_SEMICOLON_INDEX} + 1")
# STRING(SUBSTRING "${CUDA_ARCH_LIST}" ${SPLIT_START_INDEX} -1 FILTERED_CUDA_ARCH_LIST)
# set(CMAKE_CUDA_ARCHITECTURES ${FILTERED_CUDA_ARCH_LIST})

[Wanglongzhi2001]

master分支有一个被## ===================== A solution for compiling error of half caused by low cuda arch detected =======================包裹的几行代码，把那几行取消掉注释可以解决这个问题，即以下代码：

# include(FindCUDA/select_compute_arch)
# CUDA_DETECT_INSTALLED_GPUS(INSTALLED_GPU_CCS_1)
# string(STRIP "${INSTALLED_GPU_CCS_1}" INSTALLED_GPU_CCS_2)
# string(REPLACE " " "\;" INSTALLED_GPU_CCS_3 "${INSTALLED_GPU_CCS_2}")
# string(REPLACE "." "" CUDA_ARCH_LIST "${INSTALLED_GPU_CCS_3}")
# STRING(FIND "${CUDA_ARCH_LIST}" "\;" LAST_SEMICOLON_INDEX REVERSE)
# math(EXPR SPLIT_START_INDEX "${LAST_SEMICOLON_INDEX} + 1")
# STRING(SUBSTRING "${CUDA_ARCH_LIST}" ${SPLIT_START_INDEX} -1 FILTERED_CUDA_ARCH_LIST)
# set(CMAKE_CUDA_ARCHITECTURES ${FILTERED_CUDA_ARCH_LIST})

谢谢，已解决~

[daquexian]

最新的 commit 修复了这个问题~ 不取消那几行代码的注释也不会编译错误了

[Wanglongzhi2001]

最新的 commit 修复了这个问题~ 不取消那几行代码的注释也不会编译错误了

好的，我拉取下最新的代码

[daquexian]

最新的 commit 修复了这个问题~ 不取消那几行代码的注释也不会编译错误了

好的，我拉取下最新的代码

好的，有任何问题都可以继续提出！