Skip to content
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

[PHI decoupling] move norm_utils.cu.h from fluid to phi and remove norm_utils.h in fluid #48930

Merged
merged 5 commits into from
Dec 12, 2022
Merged

[PHI decoupling] move norm_utils.cu.h from fluid to phi and remove norm_utils.h in fluid #48930

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
5 commits
Select commit Hold shift + click to select a range
4f786fc
move norm_utils.cu.h from fluid to phi
huangjiyi Dec 8, 2022
073f59b
remove norm_utils.h in fluid
huangjiyi Dec 8, 2022
b01139f
fix bugs and replace mutable_data with Alloc
huangjiyi Dec 8, 2022
53bfac4
Merge branch 'develop' into decouple_norm_utils_cu_h
huangjiyi Dec 9, 2022
9cb04ba
replace mutable_data with Alloc
huangjiyi Dec 9, 2022
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
4 changes: 2 additions & 2 deletions paddle/fluid/operators/batch_norm_op.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -25,8 +25,8 @@ namespace cub = hipcub;
#endif
#include "paddle/fluid/framework/data_layout.h"
#include "paddle/fluid/operators/batch_norm_op.h"
#include "paddle/fluid/operators/norm_utils.cu.h"
#include "paddle/fluid/platform/float16.h"
#include "paddle/phi/backends/gpu/gpu_dnn.h"
#include "paddle/phi/kernels/funcs/math_function.h"

DECLARE_bool(cudnn_batchnorm_spatial_persistent);
Expand All @@ -36,7 +36,7 @@ namespace operators {

using DataLayout = phi::DataLayout;
template <typename T>
using CudnnDataType = platform::CudnnDataType<T>;
using CudnnDataType = phi::backends::gpu::CudnnDataType<T>;
template <typename T>
using BatchNormParamType = typename CudnnDataType<T>::BatchNormParamType;

Expand Down
2 changes: 1 addition & 1 deletion paddle/fluid/operators/batch_norm_op.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,8 +21,8 @@ limitations under the License. */
#include "paddle/fluid/framework/eigen.h"
#include "paddle/fluid/framework/op_registry.h"
#include "paddle/fluid/operators/layout_utils.h"
#include "paddle/fluid/operators/norm_utils.h"
#include "paddle/phi/kernels/funcs/math_function.h"
#include "paddle/phi/kernels/funcs/norm_utils.h"

namespace paddle {
namespace operators {
Expand Down
6 changes: 3 additions & 3 deletions paddle/fluid/operators/fused/fused_bn_activation_op.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,10 +21,10 @@
#include "paddle/fluid/framework/data_layout.h"
#include "paddle/fluid/operators/activation_op.h"
#include "paddle/fluid/operators/fused/fused_bn_activation_op.h"
#include "paddle/fluid/operators/norm_utils.h"
#include "paddle/fluid/platform/device/gpu/gpu_dnn.h"
#include "paddle/fluid/platform/float16.h"
#include "paddle/phi/kernels/funcs/math_function.h"
#include "paddle/phi/kernels/funcs/norm_utils.h"

DECLARE_bool(cudnn_batchnorm_spatial_persistent);

Expand Down Expand Up @@ -91,7 +91,7 @@ class FusedBatchNormActKernel<phi::GPUContext, T>

int N, C, H, W, D;
const DataLayout data_layout = DataLayout::kNHWC;
ExtractNCWHD(x_dims, data_layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x_dims, data_layout, &N, &C, &H, &W, &D);

if ((N * H * W * D) == 1) {
// Only 1 element in normalization dimension,
Expand Down Expand Up @@ -257,7 +257,7 @@ class FusedBatchNormActGradKernel<phi::GPUContext, T>
"The Input dim size should be between 2 and 5"));
int N, C, H, W, D;
const DataLayout data_layout = DataLayout::kNHWC;
ExtractNCWHD(x_dims, data_layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x_dims, data_layout, &N, &C, &H, &W, &D);

// init output
auto *d_x = ctx.Output<phi::DenseTensor>(framework::GradVarName("X"));
Expand Down
6 changes: 3 additions & 3 deletions paddle/fluid/operators/fused/fused_bn_add_activation_op.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -20,11 +20,11 @@
#include "paddle/fluid/framework/data_layout.h"
#include "paddle/fluid/operators/activation_op.h"
#include "paddle/fluid/operators/fused/fused_bn_add_activation_op.h"
#include "paddle/fluid/operators/norm_utils.h"
#include "paddle/fluid/platform/device/gpu/gpu_dnn.h"
#include "paddle/fluid/platform/float16.h"
#include "paddle/phi/common/data_type.h"
#include "paddle/phi/kernels/funcs/math_function.h"
#include "paddle/phi/kernels/funcs/norm_utils.h"

DECLARE_bool(cudnn_batchnorm_spatial_persistent);

Expand Down Expand Up @@ -85,7 +85,7 @@ class FusedBatchNormAddActKernel<phi::GPUContext, T>

int N, C, H, W, D;
const DataLayout data_layout = DataLayout::kNHWC;
ExtractNCWHD(in_dims, data_layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(in_dims, data_layout, &N, &C, &H, &W, &D);

// ------------------- cudnn descriptors ---------------------
auto handle = dev_ctx.cudnn_handle();
Expand Down Expand Up @@ -231,7 +231,7 @@ class FusedBatchNormAddActGradKernel<phi::GPUContext, T>

int N, C, H, W, D;
const DataLayout data_layout = DataLayout::kNHWC;
ExtractNCWHD(in_dims, data_layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(in_dims, data_layout, &N, &C, &H, &W, &D);

// init output
auto *d_x = ctx.Output<phi::DenseTensor>(framework::GradVarName("X"));
Expand Down
51 changes: 0 additions & 51 deletions paddle/fluid/operators/norm_utils.h
View file
Open in desktop

This file was deleted.

4 changes: 2 additions & 2 deletions paddle/fluid/operators/sync_batch_norm_op_mlu.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -72,7 +72,7 @@ class SyncBatchNormMLUKernel : public framework::OpKernel<T> {
"The Input dim size should be less than 6."));

int N, C, H, W, D;
ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);

y->mutable_data<T>(ctx.GetPlace());
mean_out->mutable_data<MPDType>(ctx.GetPlace());
Expand Down Expand Up @@ -320,7 +320,7 @@ class SyncBatchNormMLUGradKernel : public framework::OpKernel<T> {
"The Input X dim size should be less than 6."));

int N, C, H, W, D;
ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);
PADDLE_ENFORCE_EQ(scale->dims()[0],
C,
platform::errors::InvalidArgument(
Expand Down
4 changes: 2 additions & 2 deletions paddle/fluid/operators/sync_batch_norm_op_npu.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -344,7 +344,7 @@ class SyncBatchNormNPUKernel : public framework::OpKernel<T> {
x_dims.size()));

int N, C, H, W, D;
ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x_dims, layout, &N, &C, &H, &W, &D);

int x_numel = x->numel();
auto place = ctx.GetPlace();
Expand Down Expand Up @@ -598,7 +598,7 @@ class SyncBatchNormNPUGradKernel : public framework::OpKernel<T> {
}

int N, C, H, W, D;
ExtractNCWHD(x->dims(), layout, &N, &C, &H, &W, &D);
phi::funcs::ExtractNCWHD(x->dims(), layout, &N, &C, &H, &W, &D);

int x_numel = x->numel();
auto place = ctx.GetPlace();
Expand Down
20 changes: 10 additions & 10 deletions paddle/fluid/operators/norm_utils.cu.h → paddle/phi/kernels/funcs/norm_utils.cu.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -24,8 +24,7 @@ limitations under the License. */
#include <hipcub/hipcub.hpp>
namespace cub = hipcub;
#endif
#include "paddle/fluid/framework/data_layout.h"
#include "paddle/fluid/platform/device/gpu/gpu_dnn.h"
#include "paddle/phi/common/layout.h"
#include "paddle/phi/kernels/funcs/math_function.h"

#ifdef __HIPCC__
Expand All @@ -34,8 +33,8 @@ namespace cub = hipcub;
#define LAUNCH_BOUNDS(BlockDim)
#endif

namespace paddle {
namespace operators {
namespace phi {
namespace funcs {

using DataLayout = phi::DataLayout;

Expand Down Expand Up @@ -464,7 +463,8 @@ void NormDoubleGradFunctor(const DeviceContext &ctx,
const int sample_size = num / N / C;
phi::DenseTensor scale_tmp;
if (!Scale) {
scale_tmp.mutable_data<T>({C}, ctx.GetPlace());
scale_tmp.Resize({C});
ctx.template Alloc<T>(&scale_tmp);
set_constant(ctx, &scale_tmp, static_cast<T>(1));
}
const T *scale_data = Scale ? Scale->data<T>() : scale_tmp.data<T>();
Expand Down Expand Up @@ -495,7 +495,7 @@ void NormDoubleGradFunctor(const DeviceContext &ctx,
}

if (dX) {
T *dx_data = dX->mutable_data<T>(ctx.GetPlace());
T *dx_data = ctx.template Alloc<T>(dX);
set_constant(ctx, dX, static_cast<T>(0));
if (use_global_stats) {
if (data_layout == DataLayout::kNHWC) {
Expand Down Expand Up @@ -552,7 +552,7 @@ void NormDoubleGradFunctor(const DeviceContext &ctx,
}
}
if (dScale) {
T *dscale_data = dScale->mutable_data<T>(ctx.GetPlace());
T *dscale_data = ctx.template Alloc<T>(dScale);
set_constant(ctx, dScale, static_cast<T>(0));
if (use_global_stats) {
if (data_layout == DataLayout::kNHWC) {
Expand Down Expand Up @@ -605,7 +605,7 @@ void NormDoubleGradFunctor(const DeviceContext &ctx,
}
}
if (ddY) {
T *ddy_data = ddY->mutable_data<T>(ctx.GetPlace());
T *ddy_data = ctx.template Alloc<T>(ddY);
set_constant(ctx, ddY, static_cast<T>(0));
if (use_global_stats) {
if (data_layout == DataLayout::kNHWC) {
Expand Down Expand Up @@ -670,5 +670,5 @@ void NormDoubleGradFunctor(const DeviceContext &ctx,
}
}
}
} // namespace operators
} // namespace paddle
} // namespace funcs
} // namespace phi
37 changes: 18 additions & 19 deletions paddle/phi/kernels/gpu/batch_norm_grad_kernel.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -13,7 +13,6 @@
// limitations under the License.

#include "paddle/fluid/operators/layout_utils.h"
#include "paddle/fluid/operators/norm_utils.cu.h"
#include "paddle/phi/backends/gpu/gpu_context.h"
#include "paddle/phi/backends/gpu/gpu_dnn.h"
#include "paddle/phi/common/layout.h"
Expand All @@ -24,6 +23,7 @@
#include "paddle/phi/kernels/empty_kernel.h"
#include "paddle/phi/kernels/funcs/batch_norm_utils.h"
#include "paddle/phi/kernels/funcs/eigen/common.h"
#include "paddle/phi/kernels/funcs/norm_utils.cu.h"
#include "paddle/phi/kernels/funcs/norm_utils.h"
#include "paddle/phi/kernels/funcs/reduce_function.h"

Expand Down Expand Up @@ -1331,24 +1331,23 @@ void BatchNormDoubleGradKernel(
running_mean = mean.get_ptr();
running_variance = variance.get_ptr();
}
paddle::operators::NormDoubleGradFunctor<Context, T>(
ctx,
data_layout,
&x,
&scale,
&y_grad,
&saved_mean,
&saved_variance,
running_mean,
running_variance,
epsilon,
use_global_stats,
x_grad_grad.get_ptr(),
scale_grad_grad.get_ptr(),
bias_grad_grad.get_ptr(),
x_grad,
scale_grad,
y_grad_grad);
phi::funcs::NormDoubleGradFunctor<Context, T>(ctx,
data_layout,
&x,
&scale,
&y_grad,
&saved_mean,
&saved_variance,
running_mean,
running_variance,
epsilon,
use_global_stats,
x_grad_grad.get_ptr(),
scale_grad_grad.get_ptr(),
bias_grad_grad.get_ptr(),
x_grad,
scale_grad,
y_grad_grad);
}

} // namespace phi
Expand Down
1 change: 0 additions & 1 deletion paddle/phi/kernels/gpu/batch_norm_kernel.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,7 +21,6 @@ namespace cub = hipcub;
#endif

#include "paddle/fluid/operators/layout_utils.h"
#include "paddle/fluid/operators/norm_utils.cu.h"
#include "paddle/phi/backends/gpu/gpu_context.h"
#include "paddle/phi/backends/gpu/gpu_dnn.h"
#include "paddle/phi/common/layout.h"
Expand Down