Skip to content

P
Paddle

PaddlePaddle / Paddle
大约 1 年 前同步成功

通知 2299
Star 20931
Fork 5422
P
Paddle
提交 9bcd9f66 编写于 作者: C chengduo 提交者： Abhinav Arora
浏览文件
操作

fix cpplint error (#10329)

上级 55f0d840
隐藏空白更改
内联 并排
Showing with 154 addition and 120 deletion
paddle/fluid/operators/math/pooling.cc
浏览文件 @ 9bcd9f66
...@@ -11,8 +11,9 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,8 +11,9 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/math/pooling.h" #include "paddle/fluid/operators/math/pooling.h"
#include <algorithm>
#include <vector>
namespace paddle { namespace paddle {
namespace operators { namespace operators {
...@@ -27,9 +28,10 @@ template <typename PoolProcess, typename T> ...@@ -27,9 +28,10 @@ template <typename PoolProcess, typename T>
class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> { class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
PoolProcess pool_process, framework::Tensor* output) { const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* output) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
const int input_width = input.dims()[3]; const int input_width = input.dims()[3];
...@@ -63,11 +65,11 @@ class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -63,11 +65,11 @@ class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
T ele = pool_process.initial(); T ele = pool_process.initial();
for (int h = hstart; h < hend; ++h) { for (int h = hstart; h < hend; ++h) {
for (int w = wstart; w < wend; ++w) { for (int w = wstart; w < wend; ++w) {
pool_process.compute(ele, input_data[h * input_width + w]); pool_process.compute(input_data[h * input_width + w], &ele);
} }
} }
int pool_size = (hend - hstart) * (wend - wstart); int pool_size = (hend - hstart) * (wend - wstart);
pool_process.finalize(ele, (static_cast<T>(pool_size))); pool_process.finalize(static_cast<T>(pool_size), &ele);
output_data[ph * output_width + pw] = ele; output_data[ph * output_width + pw] = ele;
} }
} }
...@@ -86,13 +88,12 @@ class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -86,13 +88,12 @@ class Pool2dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
template <typename PoolProcess, class T> template <typename PoolProcess, class T>
class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(
const framework::Tensor& input, const platform::CPUDeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output, const framework::Tensor& output_grad,
const framework::Tensor& output_grad, std::vector<int>& ksize, const std::vector<int>& ksize, const std::vector<int>& strides,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& paddings, PoolProcess pool_grad_process,
PoolProcess pool_grad_process, framework::Tensor* input_grad) {
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
const int input_width = input.dims()[3]; const int input_width = input.dims()[3];
...@@ -131,8 +132,8 @@ class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -131,8 +132,8 @@ class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
input_data[h * input_width + w], input_data[h * input_width + w],
output_data[ph * output_width + pw], output_data[ph * output_width + pw],
output_grad_data[ph * output_width + pw], output_grad_data[ph * output_width + pw],
input_grad_data[h * input_width + w], static_cast<T>(scale),
static_cast<T>(scale)); input_grad_data + h * input_width + w);
} }
} }
} }
...@@ -154,12 +155,11 @@ class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -154,12 +155,11 @@ class Pool2dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
template <class T> template <class T>
class MaxPool2dGradFunctor<platform::CPUDeviceContext, T> { class MaxPool2dGradFunctor<platform::CPUDeviceContext, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(
const framework::Tensor& input, const platform::CPUDeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output, const framework::Tensor& output_grad,
const framework::Tensor& output_grad, std::vector<int>& ksize, const std::vector<int>& ksize, const std::vector<int>& strides,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& paddings, framework::Tensor* input_grad) {
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
const int input_width = input.dims()[3]; const int input_width = input.dims()[3];
...@@ -246,9 +246,10 @@ template <typename PoolProcess, class T> ...@@ -246,9 +246,10 @@ template <typename PoolProcess, class T>
class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> { class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
PoolProcess pool_process, framework::Tensor* output) { const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* output) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
const int input_height = input.dims()[3]; const int input_height = input.dims()[3];
...@@ -293,14 +294,14 @@ class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -293,14 +294,14 @@ class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
for (int h = hstart; h < hend; ++h) { for (int h = hstart; h < hend; ++h) {
for (int w = wstart; w < wend; ++w) { for (int w = wstart; w < wend; ++w) {
pool_process.compute( pool_process.compute(
ele, input_data[(d * input_height + h) * input_width + w],
input_data[(d * input_height + h) * input_width + w]); &ele);
} }
} }
} }
int pool_size = int pool_size =
(dend - dstart) * (hend - hstart) * (wend - wstart); (dend - dstart) * (hend - hstart) * (wend - wstart);
pool_process.finalize(ele, static_cast<T>(pool_size)); pool_process.finalize(static_cast<T>(pool_size), &ele);
output_data[output_idx] = ele; output_data[output_idx] = ele;
} }
} }
...@@ -320,13 +321,12 @@ class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -320,13 +321,12 @@ class Pool3dFunctor<platform::CPUDeviceContext, PoolProcess, T> {
template <typename PoolProcess, class T> template <typename PoolProcess, class T>
class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(
const framework::Tensor& input, const platform::CPUDeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output, const framework::Tensor& output_grad,
const framework::Tensor& output_grad, std::vector<int>& ksize, const std::vector<int>& ksize, const std::vector<int>& strides,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& paddings, PoolProcess pool_grad_process,
PoolProcess pool_grad_process, framework::Tensor* input_grad) {
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
const int input_height = input.dims()[3]; const int input_height = input.dims()[3];
...@@ -379,8 +379,8 @@ class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -379,8 +379,8 @@ class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
(pd * output_height + ph) * output_width + pw; (pd * output_height + ph) * output_width + pw;
pool_grad_process.compute( pool_grad_process.compute(
input_data[input_idx], output_data[output_idx], input_data[input_idx], output_data[output_idx],
output_grad_data[output_idx], output_grad_data[output_idx], static_cast<T>(scale),
input_grad_data[input_idx], static_cast<T>(scale)); input_grad_data + input_idx);
} }
} }
} }
...@@ -404,12 +404,11 @@ class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> { ...@@ -404,12 +404,11 @@ class Pool3dGradFunctor<platform::CPUDeviceContext, PoolProcess, T> {
template <class T> template <class T>
class MaxPool3dGradFunctor<platform::CPUDeviceContext, T> { class MaxPool3dGradFunctor<platform::CPUDeviceContext, T> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(
const framework::Tensor& input, const platform::CPUDeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output, const framework::Tensor& output_grad,
const framework::Tensor& output_grad, std::vector<int>& ksize, const std::vector<int>& ksize, const std::vector<int>& strides,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& paddings, framework::Tensor* input_grad) {
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
const int input_height = input.dims()[3]; const int input_height = input.dims()[3];
...@@ -510,9 +509,10 @@ template <typename T1, typename T2> ...@@ -510,9 +509,10 @@ template <typename T1, typename T2>
class MaxPool2dWithIndexFunctor<platform::CPUDeviceContext, T1, T2> { class MaxPool2dWithIndexFunctor<platform::CPUDeviceContext, T1, T2> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
framework::Tensor* output, framework::Tensor* mask) { const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
const int input_width = input.dims()[3]; const int input_width = input.dims()[3];
...@@ -576,8 +576,9 @@ class MaxPool2dWithIndexGradFunctor<platform::CPUDeviceContext, T1, T2> { ...@@ -576,8 +576,9 @@ class MaxPool2dWithIndexGradFunctor<platform::CPUDeviceContext, T1, T2> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input_grad->dims()[0]; const int batch_size = input_grad->dims()[0];
const int input_height = input_grad->dims()[2]; const int input_height = input_grad->dims()[2];
...@@ -628,9 +629,10 @@ template <typename T1, typename T2> ...@@ -628,9 +629,10 @@ template <typename T1, typename T2>
class MaxPool3dWithIndexFunctor<platform::CPUDeviceContext, T1, T2> { class MaxPool3dWithIndexFunctor<platform::CPUDeviceContext, T1, T2> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
framework::Tensor* output, framework::Tensor* mask) { const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
const int input_height = input.dims()[3]; const int input_height = input.dims()[3];
...@@ -708,8 +710,9 @@ class MaxPool3dWithIndexGradFunctor<platform::CPUDeviceContext, T1, T2> { ...@@ -708,8 +710,9 @@ class MaxPool3dWithIndexGradFunctor<platform::CPUDeviceContext, T1, T2> {
public: public:
void operator()(const platform::CPUDeviceContext& context, void operator()(const platform::CPUDeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input_grad->dims()[0]; const int batch_size = input_grad->dims()[0];
const int input_depth = input_grad->dims()[2]; const int input_depth = input_grad->dims()[2];
......
paddle/fluid/operators/math/pooling.cu
浏览文件 @ 9bcd9f66
...@@ -12,6 +12,8 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. ...@@ -12,6 +12,8 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include <algorithm>
#include <vector>
#include "paddle/fluid/operators/math/pooling.h" #include "paddle/fluid/operators/math/pooling.h"
#include "paddle/fluid/platform/cuda_primitives.h" #include "paddle/fluid/platform/cuda_primitives.h"
...@@ -47,11 +49,11 @@ __global__ void KernelPool2D(const int nthreads, const T* input_data, ...@@ -47,11 +49,11 @@ __global__ void KernelPool2D(const int nthreads, const T* input_data,
T ele = pool_process.initial(); T ele = pool_process.initial();
for (int h = hstart; h < hend; ++h) { for (int h = hstart; h < hend; ++h) {
for (int w = wstart; w < wend; ++w) { for (int w = wstart; w < wend; ++w) {
pool_process.compute(ele, input_data[h * input_width + w]); pool_process.compute(input_data[h * input_width + w], &ele);
} }
} }
int pool_size = (hend - hstart) * (wend - wstart); int pool_size = (hend - hstart) * (wend - wstart);
pool_process.finalize(ele, (static_cast<T>(pool_size))); pool_process.finalize(static_cast<T>(pool_size), &ele);
output_data[index] = ele; output_data[index] = ele;
} }
} }
...@@ -96,8 +98,8 @@ __global__ void KernelPool2DGrad( ...@@ -96,8 +98,8 @@ __global__ void KernelPool2DGrad(
int pool_size = (hend - hstart) * (wend - wstart); int pool_size = (hend - hstart) * (wend - wstart);
int output_sub_idx = ph * output_width + pw; int output_sub_idx = ph * output_width + pw;
pool_process.compute(input, output_data[output_sub_idx], pool_process.compute(input, output_data[output_sub_idx],
output_grad[output_sub_idx], gradient, output_grad[output_sub_idx],
static_cast<T>(1.0 / pool_size)); static_cast<T>(1.0 / pool_size), &gradient);
} }
} }
input_grad[index] = gradient; input_grad[index] = gradient;
...@@ -158,9 +160,10 @@ template <typename PoolProcess, typename T> ...@@ -158,9 +160,10 @@ template <typename PoolProcess, typename T>
class Pool2dFunctor<platform::CUDADeviceContext, PoolProcess, T> { class Pool2dFunctor<platform::CUDADeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
PoolProcess pool_process, framework::Tensor* output) { const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* output) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
...@@ -201,9 +204,11 @@ class Pool2dGradFunctor<platform::CUDADeviceContext, PoolProcess, T> { ...@@ -201,9 +204,11 @@ class Pool2dGradFunctor<platform::CUDADeviceContext, PoolProcess, T> {
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
PoolProcess pool_process, framework::Tensor* input_grad) { const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
...@@ -246,8 +251,10 @@ class MaxPool2dGradFunctor<platform::CUDADeviceContext, T> { ...@@ -246,8 +251,10 @@ class MaxPool2dGradFunctor<platform::CUDADeviceContext, T> {
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
...@@ -340,12 +347,12 @@ __global__ void KernelPool3D(const int nthreads, const T* input_data, ...@@ -340,12 +347,12 @@ __global__ void KernelPool3D(const int nthreads, const T* input_data,
for (int h = hstart; h < hend; ++h) { for (int h = hstart; h < hend; ++h) {
for (int w = wstart; w < wend; ++w) { for (int w = wstart; w < wend; ++w) {
pool_process.compute( pool_process.compute(
ele, input_data[(d * input_height + h) * input_width + w]); input_data[(d * input_height + h) * input_width + w], &ele);
} }
} }
} }
int pool_size = (dend - dstart) * (hend - hstart) * (wend - wstart); int pool_size = (dend - dstart) * (hend - hstart) * (wend - wstart);
pool_process.finalize(ele, static_cast<T>(pool_size)); pool_process.finalize(static_cast<T>(pool_size), &ele);
output_data[index] = ele; output_data[index] = ele;
} }
} }
...@@ -405,8 +412,8 @@ __global__ void KernelPool3DGrad( ...@@ -405,8 +412,8 @@ __global__ void KernelPool3DGrad(
int pool_size = (dend - dstart) * (hend - hstart) * (wend - wstart); int pool_size = (dend - dstart) * (hend - hstart) * (wend - wstart);
int output_sub_idx = (pd * output_height + ph) * output_width + pw; int output_sub_idx = (pd * output_height + ph) * output_width + pw;
pool_process.compute(input, output_data[output_sub_idx], pool_process.compute(input, output_data[output_sub_idx],
output_grad[output_sub_idx], gradient, output_grad[output_sub_idx],
static_cast<T>(1.0 / pool_size)); static_cast<T>(1.0 / pool_size), &gradient);
} }
} }
} }
...@@ -474,9 +481,10 @@ template <typename PoolProcess, class T> ...@@ -474,9 +481,10 @@ template <typename PoolProcess, class T>
class Pool3dFunctor<platform::CUDADeviceContext, PoolProcess, T> { class Pool3dFunctor<platform::CUDADeviceContext, PoolProcess, T> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
PoolProcess pool_process, framework::Tensor* output) { const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* output) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
...@@ -525,9 +533,11 @@ class Pool3dGradFunctor<platform::CUDADeviceContext, PoolProcess, T> { ...@@ -525,9 +533,11 @@ class Pool3dGradFunctor<platform::CUDADeviceContext, PoolProcess, T> {
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
PoolProcess pool_process, framework::Tensor* input_grad) { const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_process,
framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
...@@ -578,8 +588,10 @@ class MaxPool3dGradFunctor<platform::CUDADeviceContext, T> { ...@@ -578,8 +588,10 @@ class MaxPool3dGradFunctor<platform::CUDADeviceContext, T> {
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
...@@ -736,9 +748,10 @@ template <typename T1, typename T2> ...@@ -736,9 +748,10 @@ template <typename T1, typename T2>
class MaxPool2dWithIndexFunctor<platform::CUDADeviceContext, T1, T2> { class MaxPool2dWithIndexFunctor<platform::CUDADeviceContext, T1, T2> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
framework::Tensor* output, framework::Tensor* mask) { const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_height = input.dims()[2]; const int input_height = input.dims()[2];
...@@ -779,8 +792,9 @@ class MaxPool2dWithIndexGradFunctor<platform::CUDADeviceContext, T1, T2> { ...@@ -779,8 +792,9 @@ class MaxPool2dWithIndexGradFunctor<platform::CUDADeviceContext, T1, T2> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input_grad->dims()[0]; const int batch_size = input_grad->dims()[0];
const int input_channels = input_grad->dims()[1]; const int input_channels = input_grad->dims()[1];
...@@ -937,9 +951,10 @@ template <typename T1, typename T2> ...@@ -937,9 +951,10 @@ template <typename T1, typename T2>
class MaxPool3dWithIndexFunctor<platform::CUDADeviceContext, T1, T2> { class MaxPool3dWithIndexFunctor<platform::CUDADeviceContext, T1, T2> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& input, std::vector<int>& ksize, const framework::Tensor& input, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
framework::Tensor* output, framework::Tensor* mask) { const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask) {
const int batch_size = input.dims()[0]; const int batch_size = input.dims()[0];
const int input_channels = input.dims()[1]; const int input_channels = input.dims()[1];
const int input_depth = input.dims()[2]; const int input_depth = input.dims()[2];
...@@ -987,8 +1002,9 @@ class MaxPool3dWithIndexGradFunctor<platform::CUDADeviceContext, T1, T2> { ...@@ -987,8 +1002,9 @@ class MaxPool3dWithIndexGradFunctor<platform::CUDADeviceContext, T1, T2> {
public: public:
void operator()(const platform::CUDADeviceContext& context, void operator()(const platform::CUDADeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad) { framework::Tensor* input_grad) {
const int batch_size = input_grad->dims()[0]; const int batch_size = input_grad->dims()[0];
const int input_channels = input_grad->dims()[1]; const int input_channels = input_grad->dims()[1];
......
paddle/fluid/operators/math/pooling.h
浏览文件 @ 9bcd9f66
...@@ -13,6 +13,7 @@ See the License for the specific language governing permissions and ...@@ -13,6 +13,7 @@ See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#pragma once #pragma once
#include <vector>
#include "paddle/fluid/framework/eigen.h" #include "paddle/fluid/framework/eigen.h"
#include "paddle/fluid/framework/tensor.h" #include "paddle/fluid/framework/tensor.h"
#include "paddle/fluid/platform/device_context.h" #include "paddle/fluid/platform/device_context.h"
...@@ -23,8 +24,8 @@ namespace operators { ...@@ -23,8 +24,8 @@ namespace operators {
namespace math { namespace math {
#define FLT_MAX \ #define FLT_MAX \
__FLT_MAX__ // It might need to be placed in another file, but I'm still __FLT_MAX__ // TODO(zcd) :It might need to be placed in another file, but I'm
// wondering where to put it. // still wondering where to put it.
/* /*
* \brief Extracting simple operations from pooling. * \brief Extracting simple operations from pooling.
...@@ -40,33 +41,33 @@ template <class T> ...@@ -40,33 +41,33 @@ template <class T>
class MaxPool { class MaxPool {
public: public:
DEVICE inline T initial() { return static_cast<T>(-FLT_MAX); } DEVICE inline T initial() { return static_cast<T>(-FLT_MAX); }
DEVICE inline void compute(T& y, const T& x) { y = y > x ? y : x; } DEVICE inline void compute(const T& x, T* y) { *y = *y > x ? *y : x; }
DEVICE inline void finalize(T& y, const T& pool_field) {} DEVICE inline void finalize(const T& pool_field, T* y) {}
}; };
template <class T> template <class T>
class AvgPool { class AvgPool {
public: public:
DEVICE inline T initial() { return static_cast<T>(0); } DEVICE inline T initial() { return static_cast<T>(0); }
DEVICE inline void compute(T& y, const T& x) { y += x; } DEVICE inline void compute(const T& x, T* y) { *y += x; }
DEVICE inline void finalize(T& y, const T& pool_field) { y /= pool_field; } DEVICE inline void finalize(const T& pool_field, T* y) { *y /= pool_field; }
}; };
template <class T> template <class T>
class MaxPoolGrad { class MaxPoolGrad {
public: public:
DEVICE inline void compute(const T& x, const T& y, const T& dy, T& dx, DEVICE inline void compute(const T& x, const T& y, const T& dy, T scale,
T scale) { T* dx) {
dx += dy * (x == y); *dx += dy * (x == y);
} }
}; };
template <class T> template <class T>
class AvgPoolGrad { class AvgPoolGrad {
public: public:
DEVICE inline void compute(const T& x, const T& y, const T& dy, T& dx, DEVICE inline void compute(const T& x, const T& y, const T& dy, T scale,
T scale) { T* dx) {
dx += (scale * dy); *dx += (scale * dy);
} }
}; };
...@@ -88,8 +89,9 @@ template <typename DeviceContext, typename PoolProcess, typename T> ...@@ -88,8 +89,9 @@ template <typename DeviceContext, typename PoolProcess, typename T>
class Pool2dFunctor { class Pool2dFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
std::vector<int>& ksize, std::vector<int>& strides, const std::vector<int>& ksize,
std::vector<int>& paddings, PoolProcess pool_compute, const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_compute,
framework::Tensor* output); framework::Tensor* output);
}; };
...@@ -98,9 +100,11 @@ class Pool2dGradFunctor { ...@@ -98,9 +100,11 @@ class Pool2dGradFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
PoolProcess pool_compute, framework::Tensor* input_grad); const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_compute,
framework::Tensor* input_grad);
}; };
template <typename DeviceContext, class T> template <typename DeviceContext, class T>
...@@ -108,8 +112,10 @@ class MaxPool2dGradFunctor { ...@@ -108,8 +112,10 @@ class MaxPool2dGradFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad); framework::Tensor* input_grad);
}; };
...@@ -117,8 +123,9 @@ template <typename DeviceContext, typename PoolProcess, typename T> ...@@ -117,8 +123,9 @@ template <typename DeviceContext, typename PoolProcess, typename T>
class Pool3dFunctor { class Pool3dFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
std::vector<int>& ksize, std::vector<int>& strides, const std::vector<int>& ksize,
std::vector<int>& paddings, PoolProcess pool_compute, const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_compute,
framework::Tensor* output); framework::Tensor* output);
}; };
...@@ -127,9 +134,11 @@ class Pool3dGradFunctor { ...@@ -127,9 +134,11 @@ class Pool3dGradFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
PoolProcess pool_compute, framework::Tensor* input_grad); const std::vector<int>& strides,
const std::vector<int>& paddings, PoolProcess pool_compute,
framework::Tensor* input_grad);
}; };
template <typename DeviceContext, class T> template <typename DeviceContext, class T>
...@@ -137,8 +146,10 @@ class MaxPool3dGradFunctor { ...@@ -137,8 +146,10 @@ class MaxPool3dGradFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
const framework::Tensor& output, const framework::Tensor& output,
const framework::Tensor& output_grad, std::vector<int>& ksize, const framework::Tensor& output_grad,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& ksize,
const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad); framework::Tensor* input_grad);
}; };
...@@ -153,8 +164,9 @@ template <typename DeviceContext, typename T1, typename T2> ...@@ -153,8 +164,9 @@ template <typename DeviceContext, typename T1, typename T2>
class MaxPool2dWithIndexFunctor { class MaxPool2dWithIndexFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
std::vector<int>& ksize, std::vector<int>& strides, const std::vector<int>& ksize,
std::vector<int>& paddings, framework::Tensor* output, const std::vector<int>& strides,
const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask); framework::Tensor* mask);
}; };
...@@ -163,8 +175,9 @@ class MaxPool2dWithIndexGradFunctor { ...@@ -163,8 +175,9 @@ class MaxPool2dWithIndexGradFunctor {
public: public:
void operator()(const DeviceContext& context, void operator()(const DeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad); framework::Tensor* input_grad);
}; };
...@@ -172,8 +185,9 @@ template <typename DeviceContext, typename T1, typename T2> ...@@ -172,8 +185,9 @@ template <typename DeviceContext, typename T1, typename T2>
class MaxPool3dWithIndexFunctor { class MaxPool3dWithIndexFunctor {
public: public:
void operator()(const DeviceContext& context, const framework::Tensor& input, void operator()(const DeviceContext& context, const framework::Tensor& input,
std::vector<int>& ksize, std::vector<int>& strides, const std::vector<int>& ksize,
std::vector<int>& paddings, framework::Tensor* output, const std::vector<int>& strides,
const std::vector<int>& paddings, framework::Tensor* output,
framework::Tensor* mask); framework::Tensor* mask);
}; };
...@@ -182,8 +196,9 @@ class MaxPool3dWithIndexGradFunctor { ...@@ -182,8 +196,9 @@ class MaxPool3dWithIndexGradFunctor {
public: public:
void operator()(const DeviceContext& context, void operator()(const DeviceContext& context,
const framework::Tensor& output_grad, const framework::Tensor& output_grad,
const framework::Tensor& mask, std::vector<int>& ksize, const framework::Tensor& mask, const std::vector<int>& ksize,
std::vector<int>& strides, std::vector<int>& paddings, const std::vector<int>& strides,
const std::vector<int>& paddings,
framework::Tensor* input_grad); framework::Tensor* input_grad);
}; };
......
Markdown is supported
0% 或 .
You are about to add 0 people to the discussion. Proceed with caution.
先完成此消息的编辑！
想要评论请 注册 或