From b7e120d264a33d97bb7d946d6197edc488a0976c Mon Sep 17 00:00:00 2001 From: huangjiyi <43315610+huangjiyi@users.noreply.github.com> Date: Thu, 17 Nov 2022 11:38:50 +0800 Subject: [PATCH] rm "paddle/phi/kernels/gpu/batch_norm_utils.h" in phi (#48057) --- .../phi/kernels/cpu/batch_norm_grad_kernel.cc | 2 +- .../phi/kernels/gpu/batch_norm_grad_kernel.cu | 2 +- paddle/phi/kernels/gpu/batch_norm_kernel.cu | 2 +- paddle/phi/kernels/gpu/batch_norm_utils.h | 142 ------------------ 4 files changed, 3 insertions(+), 145 deletions(-) delete mode 100644 paddle/phi/kernels/gpu/batch_norm_utils.h diff --git a/paddle/phi/kernels/cpu/batch_norm_grad_kernel.cc b/paddle/phi/kernels/cpu/batch_norm_grad_kernel.cc index f2054d4d396..efd55dee88c 100644 --- a/paddle/phi/kernels/cpu/batch_norm_grad_kernel.cc +++ b/paddle/phi/kernels/cpu/batch_norm_grad_kernel.cc @@ -16,9 +16,9 @@ #include "paddle/phi/backends/cpu/cpu_context.h" #include "paddle/phi/core/kernel_registry.h" #include "paddle/phi/kernels/batch_norm_kernel.h" +#include "paddle/phi/kernels/funcs/batch_norm_utils.h" #include "paddle/phi/kernels/funcs/eigen/common.h" #include "paddle/phi/kernels/funcs/math_function.h" -#include "paddle/phi/kernels/gpu/batch_norm_utils.h" namespace phi { diff --git a/paddle/phi/kernels/gpu/batch_norm_grad_kernel.cu b/paddle/phi/kernels/gpu/batch_norm_grad_kernel.cu index 8d072368633..e6c681588e4 100644 --- a/paddle/phi/kernels/gpu/batch_norm_grad_kernel.cu +++ b/paddle/phi/kernels/gpu/batch_norm_grad_kernel.cu @@ -22,10 +22,10 @@ #include "paddle/phi/core/kernel_registry.h" #include "paddle/phi/kernels/batch_norm_kernel.h" #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.h" #include "paddle/phi/kernels/funcs/reduce_function.h" -#include "paddle/phi/kernels/gpu/batch_norm_utils.h" #ifdef __HIPCC__ #define LAUNCH_BOUNDS(BlockDim) __launch_bounds__(BlockDim) diff --git a/paddle/phi/kernels/gpu/batch_norm_kernel.cu b/paddle/phi/kernels/gpu/batch_norm_kernel.cu index 7b553db274d..44fe99046e1 100644 --- a/paddle/phi/kernels/gpu/batch_norm_kernel.cu +++ b/paddle/phi/kernels/gpu/batch_norm_kernel.cu @@ -29,10 +29,10 @@ namespace cub = hipcub; #include "paddle/phi/common/layout.h" #include "paddle/phi/core/kernel_registry.h" #include "paddle/phi/kernels/batch_norm_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.h" #include "paddle/phi/kernels/funcs/reduce_function.h" -#include "paddle/phi/kernels/gpu/batch_norm_utils.h" #ifdef __HIPCC__ #define LAUNCH_BOUNDS(BlockDim) __launch_bounds__(BlockDim) diff --git a/paddle/phi/kernels/gpu/batch_norm_utils.h b/paddle/phi/kernels/gpu/batch_norm_utils.h deleted file mode 100644 index c9c62026edf..00000000000 --- a/paddle/phi/kernels/gpu/batch_norm_utils.h +++ /dev/null @@ -1,142 +0,0 @@ -// Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved. -// -// Licensed under the Apache License, Version 2.0 (the "License"); -// you may not use this file except in compliance with the License. -// You may obtain a copy of the License at -// -// http://www.apache.org/licenses/LICENSE-2.0 -// -// Unless required by applicable law or agreed to in writing, software -// distributed under the License is distributed on an "AS IS" BASIS, -// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -// See the License for the specific language governing permissions and -// limitations under the License. - -#pragma once - -#include "paddle/phi/kernels/funcs/math_function.h" - -namespace phi { - -using Tensor = DenseTensor; - -template -inline void ResizeToChannelFirst(const DeviceContext& context, - const Tensor* input, - Tensor* transformed_input) { - int dim = input->dims().size() - 2; - if (dim == 3) { - // input - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[4]; - in_dims_vec[2] = input->dims()[1]; - in_dims_vec[3] = input->dims()[2]; - in_dims_vec[4] = input->dims()[3]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - - } else if (dim == 2) { - // input - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[3]; - in_dims_vec[2] = input->dims()[1]; - in_dims_vec[3] = input->dims()[2]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - } else if (dim == 1) { - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[2]; - in_dims_vec[2] = input->dims()[1]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - } -} - -template -inline void ResizeToChannelLast(const DeviceContext& context, - const Tensor* input, - Tensor* transformed_input) { - int dim = input->dims().size() - 2; - if (dim == 3) { - // input - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[2]; - in_dims_vec[2] = input->dims()[3]; - in_dims_vec[3] = input->dims()[4]; - in_dims_vec[4] = input->dims()[1]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - - } else if (dim == 2) { - // input - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[2]; - in_dims_vec[2] = input->dims()[3]; - in_dims_vec[3] = input->dims()[1]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - } else if (dim == 1) { - transformed_input->Resize(input->dims()); - - auto in_dims_vec = phi::vectorize(input->dims()); - in_dims_vec[1] = input->dims()[2]; - in_dims_vec[2] = input->dims()[1]; - transformed_input->Resize(phi::make_ddim(in_dims_vec)); - context.template Alloc(transformed_input); - } -} - -template -inline void TransToChannelFirst(const DeviceContext& context, - const Tensor* input, - Tensor* transformed_input) { - VLOG(5) << "Why am I called?"; - int dim = input->dims().size() - 2; - if (dim == 3) { - std::vector axis{0, 4, 1, 2, 3}; - funcs::Transpose trans5; - trans5(context, *input, transformed_input, axis); - - } else if (dim == 2) { - std::vector axis{0, 3, 1, 2}; - funcs::Transpose trans4; - trans4(context, *input, transformed_input, axis); - } else if (dim == 1) { - std::vector axis{0, 2, 1}; - funcs::Transpose trans3; - trans3(context, *input, transformed_input, axis); - } -} - -template -inline void TransToChannelLast(const DeviceContext& context, - const Tensor* input, - Tensor* transformed_input) { - int dim = input->dims().size() - 2; - if (dim == 3) { - std::vector axis{0, 2, 3, 4, 1}; - funcs::Transpose trans5; - trans5(context, *input, transformed_input, axis); - - } else if (dim == 2) { - std::vector axis{0, 2, 3, 1}; - funcs::Transpose trans4; - trans4(context, *input, transformed_input, axis); - } else if (dim == 1) { - std::vector axis{0, 2, 1}; - funcs::Transpose trans3; - trans3(context, *input, transformed_input, axis); - } -} - -} // namespace phi -- GitLab