Update Paddle to 1223 and fix dense_tensor_iterator (#2299)

* add fused_rms_norm_ext and fused_rms_norm_ext_grad

Co-authored-by: duqimeng <77875733+duqimeng@users.noreply.github.com>

Author: YqGe585

Paddle

@@ -118,6 +118,7 @@ file(
   ${PADDLE_SOURCE_DIR}/paddle/phi/core/platform/device/gpu/gpu_info.cc
   # kernels/funcs
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/*.cu
+  ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/dense_tensor_iterator.cc
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/math/*.cu
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/eigen/*.cu
   # cudnn/cublas

backends/iluvatar_gpu/CMakeLists.txt

@@ -81,8 +81,16 @@ void RMSLnBwd(const Context &ctx,
 
 }  // namespace phi
 
-PD_REGISTER_PLUGIN_KERNEL(
-    fused_rms_norm_ext, iluvatar_gpu, ALL_LAYOUT, phi::RMSLnFwd, float) {}
+PD_REGISTER_PLUGIN_KERNEL(fused_rms_norm_ext,
+                          iluvatar_gpu,
+                          ALL_LAYOUT,
+                          phi::RMSLnFwd,
+                          float,
+                          phi::bfloat16) {}
 
-PD_REGISTER_PLUGIN_KERNEL(
-    fused_rms_norm_ext_grad, iluvatar_gpu, ALL_LAYOUT, phi::RMSLnBwd, float) {}
+PD_REGISTER_PLUGIN_KERNEL(fused_rms_norm_ext_grad,
+                          iluvatar_gpu,
+                          ALL_LAYOUT,
+                          phi::RMSLnBwd,
+                          float,
+                          phi::bfloat16) {}

backends/iluvatar_gpu/kernels/ernie_core/layer_norm_cuda_kernel.cu

@@ -727,7 +727,7 @@ index 77e3537124..8f6022bc76 100644
  template <typename T>
  struct SumOp {
 diff --git a/paddle/phi/kernels/fusion/gpu/fused_layernorm_residual_dropout_bias.h b/paddle/phi/kernels/fusion/gpu/fused_layernorm_residual_dropout_bias.h
-index 9d4bb18d55..78bf0ad1b9 100644
+index a28047c624..30832164f4 100644
 --- a/paddle/phi/kernels/fusion/gpu/fused_layernorm_residual_dropout_bias.h
 +++ b/paddle/phi/kernels/fusion/gpu/fused_layernorm_residual_dropout_bias.h
 @@ -24,11 +24,7 @@ namespace fusion {
@@ -754,7 +754,7 @@ index 9d4bb18d55..78bf0ad1b9 100644
 -  __shared__ U shared_var[32];
 -#endif
 
-   phi::funcs::ReluFunctor<T> relu;
+   funcs::ReluFunctor<T> relu;
    U mean_val = 0;
 @@ -352,13 +343,8 @@ __global__ void FusedLayernormResidualDropoutBiasInfer(
 
@@ -768,7 +768,7 @@ index 9d4bb18d55..78bf0ad1b9 100644
 -  __shared__ U shared_var[32];
 -#endif
 
-   phi::funcs::ReluFunctor<T> relu;
+   funcs::ReluFunctor<T> relu;
    U mean_val = 0;
 @@ -638,9 +624,6 @@ __global__ __launch_bounds__(THREADS_PER_CTA) void fused_fast_ln_fwd_kernel(
          RandVec<VecSize>(&state, rand);
@@ -780,6 +780,7 @@ index 9d4bb18d55..78bf0ad1b9 100644
            mask_vec[it][jt] = static_cast<MaskType>(rand[jt] >= dropout_prob);
          }
        }
+
 diff --git a/paddle/phi/kernels/gpu/elementwise_grad.h b/paddle/phi/kernels/gpu/elementwise_grad.h
 index 411ee4510c..36c2f8fba7 100644
 --- a/paddle/phi/kernels/gpu/elementwise_grad.h

backends/iluvatar_gpu/patches/paddle-corex.patch

@@ -556,3 +556,4 @@ test_rms_norm_op.py
 test_batched_gemm.py
 test_match_matrix_tensor_op.py
 test_tensor.py
+test_rms_norm.py

backends/iluvatar_gpu/tests/disabled_test.txt

@@ -121,6 +121,7 @@ file(
   ${PADDLE_SOURCE_DIR}/paddle/phi/backends/dynload/cusparse.cc
   # kernels/Funcs
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/*.cu
+  ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/dense_tensor_iterator.cc
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/math/*.cu
   ${PADDLE_SOURCE_DIR}/paddle/phi/kernels/funcs/eigen/*.cu
   # kernels/gpu

backends/metax_gpu/CMakeLists.txt

@@ -0,0 +1 @@
+../../../../../../../Paddle/paddle/phi/backends/dynload/cuda_driver.h

backends/metax_gpu/headers/paddle/phi/backends/dynload/cuda_driver.h

@@ -0,0 +1 @@
+../../../../../../../Paddle/paddle/phi/backends/dynload/cudnn_frontend.h

backends/metax_gpu/headers/paddle/phi/backends/dynload/cudnn_frontend.h

@@ -0,0 +1 @@
+../../../../../../../Paddle/paddle/phi/backends/dynload/curand.h

backends/metax_gpu/headers/paddle/phi/backends/dynload/curand.h

@@ -0,0 +1 @@
+../../../../../../../Paddle/paddle/phi/backends/dynload/cusparseLt.h