Skip to content

optimize using float16 in kernels #78644

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
luotao1 merged 3 commits into from
Apr 17, 2026
Merged

optimize using float16 in kernels #78644

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
3 commits
Select commit Hold shift + click to select a range
f6799b9
optimize using float16 in kernels
co63oc Apr 12, 2026
97aa78f
Merge branch 'develop' into n320
co63oc Apr 14, 2026
b491d9a
Merge branch 'develop' into n320
co63oc Apr 16, 2026
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
1 change: 0 additions & 1 deletion paddle/phi/kernels/funcs/adam_functors.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -27,7 +27,6 @@

namespace phi {
namespace funcs {
using float16 = dtype::float16;

#ifdef PADDLE_WITH_XPU

Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/funcs/blas/blas_impl.cu.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -317,8 +317,6 @@ struct CUBlas<double> {

template <>
struct CUBlas<phi::float16> {
using float16 = phi::float16;

static void GEMM(cublasHandle_t handle,
cublasOperation_t transa,
cublasOperation_t transb,
Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/funcs/blas/blas_impl.hip.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -181,8 +181,6 @@ struct CUBlas<double> {

template <>
struct CUBlas<phi::float16> {
using float16 = phi::float16;

static void GEMM(rocblas_handle handle,
rocblas_operation transa,
rocblas_operation transb,
Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/funcs/fc_functor.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -28,8 +28,6 @@ limitations under the License. */
namespace phi {
namespace funcs {

using float16 = phi::float16;

template <typename T>
struct FcTypeTraits;

Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/funcs/math_function.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -46,8 +46,6 @@ limitations under the License. */

namespace phi::funcs {

using float16 = phi::float16;

template struct SetConstant<CPUContext, phi::float8_e4m3fn>;
template struct SetConstant<CPUContext, phi::float8_e5m2>;
template struct SetConstant<CPUContext, phi::float16>;
Expand Down
3 changes: 0 additions & 3 deletions paddle/phi/kernels/funcs/math_function.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -123,9 +123,6 @@ void BatchTranspose(T* output,
output, input, batch, m, n, swizzle);
}

using float16 = phi::float16;
using bfloat16 = phi::bfloat16;

template void BatchTranspose(float16* output,
const float16* input,
int64_t batch,
Expand Down
1 change: 0 additions & 1 deletion paddle/phi/kernels/funcs/segment_pooling.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -142,7 +142,6 @@ class SegmentPoolGradFunctor<CPUContext, T, IndexT> {
};

using CPU = CPUContext;
using float16 = phi::float16;
template class SegmentPoolFunctor<CPU, float, int>;
template class SegmentPoolFunctor<CPU, float, int64_t>;
template class SegmentPoolFunctor<CPU, double, int>;
Expand Down
1 change: 0 additions & 1 deletion paddle/phi/kernels/funcs/segment_pooling.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -440,7 +440,6 @@ class SegmentPoolGradFunctor<GPUContext, T, IndexT> {
};

using GPU = GPUContext;
using float16 = phi::float16;
template class SegmentPoolFunctor<GPU, float, int>;
template class SegmentPoolFunctor<GPU, float, int64_t>;
template class SegmentPoolFunctor<GPU, double, int>;
Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/fusion/gpu/fused_fc_elementwise_layernorm_kernel.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -31,8 +31,6 @@
namespace phi {
namespace fusion {

using float16 = phi::float16;

template <typename T>
static __device__ __forceinline__ T Relu(T x) {
return static_cast<T>(fmaxf(0.f, x));
Expand Down
2 changes: 0 additions & 2 deletions paddle/phi/kernels/fusion/gpu/fused_multi_transformer_op.cu.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -37,8 +37,6 @@ namespace fusion {

namespace { // NOLINT

using float16 = phi::float16;

#define MMHA_USE_FP32_ACUM_FOR_LOGITS
#define MMHA_USE_FP32_ACUM_FOR_OUT
#define MMHA_USE_FP32_ACUM_FOR_FMA
Expand Down
Loading