conv_op.cc 15.1 KB
Newer Older
1
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserved.
C
chengduoZH 已提交
2

L
Luo Tao 已提交
3 4 5
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
C
chengduoZH 已提交
6

L
Luo Tao 已提交
7
    http://www.apache.org/licenses/LICENSE-2.0
C
chengduoZH 已提交
8

L
Luo Tao 已提交
9 10 11 12 13
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. */
C
chengduoZH 已提交
14

Y
Yi Wang 已提交
15
#include "paddle/fluid/operators/conv_op.h"
C
chengduoZH 已提交
16 17 18 19

namespace paddle {
namespace operators {

C
chengduoZH 已提交
20
void ConvOp::InferShape(framework::InferShapeContext* ctx) const {
C
chengduoZH 已提交
21
  PADDLE_ENFORCE(ctx->HasInput("Input"),
C
chengduoZH 已提交
22
                 "Input(Input) of ConvOp should not be null.");
C
chengduoZH 已提交
23
  PADDLE_ENFORCE(ctx->HasInput("Filter"),
C
chengduoZH 已提交
24
                 "Input(Filter) of ConvOp should not be null.");
C
chengduoZH 已提交
25
  PADDLE_ENFORCE(ctx->HasOutput("Output"),
C
chengduoZH 已提交
26
                 "Output(Output) of ConvOp should not be null.");
C
chengduoZH 已提交
27 28 29 30 31 32

  auto in_dims = ctx->GetInputDim("Input");
  auto filter_dims = ctx->GetInputDim("Filter");
  std::vector<int> strides = ctx->Attrs().Get<std::vector<int>>("strides");
  std::vector<int> paddings = ctx->Attrs().Get<std::vector<int>>("paddings");
  int groups = ctx->Attrs().Get<int>("groups");
C
chengduoZH 已提交
33
  std::vector<int> dilations = ctx->Attrs().Get<std::vector<int>>("dilations");
C
chengduoZH 已提交
34

C
chengduoZH 已提交
35 36
  PADDLE_ENFORCE(in_dims.size() == 4 || in_dims.size() == 5,
                 "Conv intput should be 4-D or 5-D tensor.");
C
chengduoZH 已提交
37 38 39 40 41 42 43 44 45
  PADDLE_ENFORCE_EQ(
      in_dims.size(), filter_dims.size(),
      "Conv input dimension and filter dimension should be the same.");
  PADDLE_ENFORCE(
      in_dims.size() - strides.size() == 2U,
      "Conv input dimension and strides dimension should be consistent.");
  PADDLE_ENFORCE_EQ(
      paddings.size(), strides.size(),
      "Conv paddings dimension and Conv strides dimension should be the same.");
F
fengjiayi 已提交
46

Y
Yang Yu 已提交
47
  PADDLE_ENFORCE_EQ(in_dims[1], filter_dims[1] * groups,
C
chengduoZH 已提交
48
                    "The number of input channels should be equal to filter "
C
chengduoZH 已提交
49
                    "channels * groups.");
F
fengjiayi 已提交
50

C
chengduoZH 已提交
51
  PADDLE_ENFORCE_EQ(
Y
Yang Yu 已提交
52
      filter_dims[0] % groups, 0,
C
chengduoZH 已提交
53 54 55
      "The number of output channels should be divided by groups.");

  std::vector<int64_t> output_shape({in_dims[0], filter_dims[0]});
C
chengduoZH 已提交
56
  for (size_t i = 0; i < strides.size(); ++i) {
Y
Yang Yang 已提交
57 58 59
    output_shape.push_back(ConvOutputSize(in_dims[i + 2], filter_dims[i + 2],
                                          dilations[i], paddings[i],
                                          strides[i]));
C
chengduoZH 已提交
60
  }
61
  ctx->SetOutputDim("Output", framework::make_ddim(output_shape));
Y
Yang Yu 已提交
62
  ctx->ShareLoD("Input", "Output");
C
chengduoZH 已提交
63 64
}

65 66 67
framework::OpKernelType ConvOp::GetExpectedKernelType(
    const framework::ExecutionContext& ctx) const {
  bool use_cudnn = ctx.Attr<bool>("use_cudnn");
C
chengduoZH 已提交
68
  use_cudnn &= platform::is_gpu_place(ctx.GetPlace());
C
chengduoZH 已提交
69 70 71 72 73 74
#ifdef PADDLE_WITH_CUDA
  if (platform::is_gpu_place(ctx.GetPlace())) {
    auto& dev_ctx = ctx.template device_context<platform::CUDADeviceContext>();
    use_cudnn &= dev_ctx.cudnn_handle() != nullptr;
  }
#endif
75 76 77 78 79 80 81 82 83 84 85 86 87 88
  framework::LibraryType library_;
  if (use_cudnn) {
    library_ = framework::LibraryType::kCUDNN;
  } else {
    library_ = framework::LibraryType::kPlain;
  }

  std::string data_format = ctx.Attr<std::string>("data_format");
  framework::DataLayout layout_ = framework::StringToDataLayout(data_format);
  return framework::OpKernelType(
      framework::ToDataType(ctx.Input<Tensor>("Input")->type()), ctx.GetPlace(),
      layout_, library_);
}

89
Conv2DOpMaker::Conv2DOpMaker(OpProto* proto, OpAttrChecker* op_checker)
C
chengduoZH 已提交
90 91 92
    : OpProtoAndCheckerMaker(proto, op_checker) {
  AddInput(
      "Input",
C
fix doc  
chengduoZH 已提交
93 94 95 96
      "(Tensor) The input tensor of convolution operator. "
      "The format of input tensor is NCHW, where N is batch size, C is the "
      "number of channels, H is the height of the feature, "
      "and W is the width of the feature.");
C
chengduoZH 已提交
97
  AddInput("Filter",
C
fix doc  
chengduoZH 已提交
98
           "(Tensor) The filter tensor of convolution operator. "
C
chengduoZH 已提交
99 100
           "The format of the filter tensor is MCHW, where M is the number of "
           "output image channels, C is the number of input image channels, "
C
fix doc  
chengduoZH 已提交
101 102
           "H is the height of the filter, and W is the width of the filter. "
           "If the groups attribute is greater than 1, C equals the number of "
C
chengduoZH 已提交
103 104
           "input image channels divided by the groups.");
  AddOutput("Output",
C
fix doc  
chengduoZH 已提交
105 106
            "(Tensor) The output tensor of convolution operator. "
            "The format of output tensor is also NCHW.");
C
chengduoZH 已提交
107 108 109 110
  AddAttr<std::vector<int>>("strides",
                            "(vector<int> default:{1, 1}), the "
                            "strides(h_stride, w_stride) of "
                            "convolution operator.")
C
chengduoZH 已提交
111
      .SetDefault({1, 1});
C
chengduoZH 已提交
112 113 114 115
  AddAttr<std::vector<int>>("paddings",
                            "(vector<int> default:{0, 0}), the "
                            "paddings(h_pad, w_pad) of "
                            "convolution operator.")
C
chengduoZH 已提交
116 117 118
      .SetDefault({0, 0});
  AddAttr<int>(
      "groups",
C
chengduoZH 已提交
119
      "(int default:1), the groups number of the convolution operator. "
C
fix doc  
chengduoZH 已提交
120 121 122 123
      "According to grouped convolution in Alex Krizhevsky's Deep CNN paper: "
      "when group=2, the first half of the filters is only connected to the "
      "first half of the input channels, while the second half of the filters "
      "is only connected to the second half of the input channels.")
C
chengduoZH 已提交
124
      .SetDefault(1);
C
chengduoZH 已提交
125
  AddAttr<std::vector<int>>("dilations",
C
chengduoZH 已提交
126 127
                            "(vector<int> default:{1, 1}), the "
                            "dilations(h_dilation, w_dilation) of "
C
chengduoZH 已提交
128
                            "convolution operator.")
C
chengduoZH 已提交
129
      .SetDefault({1, 1});
130 131 132 133 134 135 136 137 138 139 140 141 142 143 144 145 146 147 148 149
  AddAttr<bool>(
      "use_cudnn",
      "(bool, default false) Only used in cudnn kernel, need install cudnn")
      .SetDefault(false);
  AddAttr<std::string>(
      "data_format",
      "(string, default NCHW) Only used in "
      "An optional string from: \"NHWC\", \"NCHW\". "
      "Defaults to \"NHWC\". Specify the data format of the output data, "
      "the input will be transformed automatically. ")
      .SetDefault("AnyLayout");
  // TODO(dzhwinter): need to registered layout transform function
  AddAttr<int>("workspace_size_MB",
               "Only used in cudnn kernel. Need set use_cudnn to true."
               "workspace size for cudnn, in MB, "
               "workspace is a section of GPU memory which will be "
               "allocated/freed each time the operator runs, larger "
               "workspace size can increase performance but also requires "
               "better hardware. This size should be chosen carefully.")
      .SetDefault(4096);
C
chengduoZH 已提交
150
  AddComment(R"DOC(
C
fix doc  
chengduoZH 已提交
151 152
Convolution Operator.

C
chengduoZH 已提交
153
The convolution operation calculates the output based on the input, filter
C
chengduoZH 已提交
154
and strides, paddings, dilations, groups parameters. The size of each dimension of the
C
chengduoZH 已提交
155
parameters is checked in the infer-shape.
C
chengduoZH 已提交
156
Input(Input) and Output(Output) are in NCHW format. Where N is batch
C
fix doc  
chengduoZH 已提交
157
size, C is the number of channels, H is the height of the feature, and W is
C
chengduoZH 已提交
158 159 160 161 162 163
the width of the feature.
Filters(Input) is MCHW format. Where M is the number of output image channels, C is
the number of input image channels, H is the height of the filter, and W
is the width of the filter.
Parameters(strides, paddings, dilations) are two elements. These two elements represent
height and width, respectively.
C
chengduoZH 已提交
164 165 166 167
The input(X) size and output(Out) size may be different.

Example:
  Input:
C
chengduoZH 已提交
168 169
       Input shape: $(N, C_{in}, H_{in}, W_{in})$
       Filter shape: $(C_{out}, C_{in}, H_f, W_f)$
C
chengduoZH 已提交
170
  Output:
C
chengduoZH 已提交
171 172 173 174 175 176
       Output shape: $(N, C_{out}, H_{out}, W_{out})$
  Where
$$
       H_{out}= \frac{(H_{in} + 2 * paddings[0] - (dilations[0] * (H_f - 1) + 1))}{strides[0]}+ 1 \\
       W_{out}= \frac{(W_{in} + 2 * paddings[1] - (dilations[1] * (W_f - 1) + 1))}{strides[1]}+ 1
$$
C
chengduoZH 已提交
177
)DOC");
C
chengduoZH 已提交
178 179
}

180
Conv3DOpMaker::Conv3DOpMaker(OpProto* proto, OpAttrChecker* op_checker)
C
chengduoZH 已提交
181 182 183
    : OpProtoAndCheckerMaker(proto, op_checker) {
  AddInput(
      "Input",
C
fix doc  
chengduoZH 已提交
184
      "(Tensor) The input tensor of convolution operator. "
C
chengduoZH 已提交
185
      "The format of input tensor is NCDHW. Where N is batch size, C is the "
C
fix doc  
chengduoZH 已提交
186 187 188
      "number of channels, D is the depth of the feature, H is the height of "
      "the feature, "
      "and W is the width of the feature.");
C
chengduoZH 已提交
189
  AddInput("Filter",
C
fix doc  
chengduoZH 已提交
190
           "(Tensor) The filter tensor of convolution operator. "
C
chengduoZH 已提交
191 192
           "The format of the filter tensor is MCDHW, where M is the number of "
           "output image channels, C is the number of input image channels, "
C
fix doc  
chengduoZH 已提交
193 194 195
           "D is the depth of the filter, H is the height of the filter, and W "
           "is the width of the filter."
           "If the groups attribute is greater than 1, C equals the number of "
C
chengduoZH 已提交
196 197
           "input image channels divided by the groups.");
  AddOutput("Output",
C
fix doc  
chengduoZH 已提交
198
            "(Tensor) The output tensor of convolution operator."
C
chengduoZH 已提交
199
            "The format of output tensor is also NCDHW.");
C
chengduoZH 已提交
200 201 202 203
  AddAttr<std::vector<int>>("strides",
                            "(vector<int>, default:{1, 1, 1}), the "
                            "strides(d_stride, h_stride, w_stride) of "
                            "convolution operator.")
C
chengduoZH 已提交
204
      .SetDefault({1, 1, 1});
C
chengduoZH 已提交
205 206 207 208
  AddAttr<std::vector<int>>("paddings",
                            "(vector<int>, default:{0, 0, 0}), the "
                            "paddings(d_pad, h_pad, w_pad) of convolution "
                            "operator.")
C
chengduoZH 已提交
209 210 211
      .SetDefault({0, 0, 0});
  AddAttr<int>(
      "groups",
C
chengduoZH 已提交
212
      "(int default:1), the groups number of the convolution operator. "
C
fix doc  
chengduoZH 已提交
213 214 215 216
      "According to grouped convolution in Alex Krizhevsky's Deep CNN paper: "
      "when group=2, the first half of the filters is only connected to the "
      "first half of the input channels, while the second half of the filters "
      "is only connected to the second half of the input channels.")
C
chengduoZH 已提交
217
      .SetDefault(1);
C
chengduoZH 已提交
218
  AddAttr<std::vector<int>>("dilations",
C
chengduoZH 已提交
219 220
                            "(vector<int> default:{1, 1, 1}), the "
                            "dilations(d_dilation, h_dilation, w_dilation) of "
C
chengduoZH 已提交
221
                            "convolution operator.")
C
chengduoZH 已提交
222
      .SetDefault({1, 1, 1});
223 224 225 226 227 228 229 230 231 232 233 234 235 236 237 238 239 240 241
  AddAttr<bool>(
      "use_cudnn",
      "(bool, default false) Only used in cudnn kernel, need install cudnn")
      .SetDefault(false);
  AddAttr<std::string>(
      "data_format",
      "(string, default NCHW) Only used in "
      "An optional string from: \"NHWC\", \"NCHW\". "
      "Defaults to \"NHWC\". Specify the data format of the output data, "
      "the input will be transformed automatically. ")
      .SetDefault("AnyLayout");
  // TODO(dzhwinter): need to registered layout transform function
  AddAttr<int>("workspace_size_MB",
               "Only used in cudnn kernel. workspace size for cudnn, in MB, "
               "workspace is a section of GPU memory which will be "
               "allocated/freed each time the operator runs, larger "
               "workspace size can increase performance but also requires "
               "better hardware. This size should be chosen carefully.")
      .SetDefault(4096);
C
fix doc  
chengduoZH 已提交
242

C
chengduoZH 已提交
243
  AddComment(R"DOC(
C
fix doc  
chengduoZH 已提交
244 245
Convolution3D Operator.

C
chengduoZH 已提交
246
The convolution operation calculates the output based on the input, filter
C
chengduoZH 已提交
247
and strides, paddings, dilations, groups parameters. The size of each dimension of the
C
chengduoZH 已提交
248
parameters is checked in the infer-shape.
C
chengduoZH 已提交
249
Input(Input) and output(Output) are in NCDHW format, where N is batch
C
fix doc  
chengduoZH 已提交
250
size, C is the number of channels,D is the depth of the feature, H is the height of
C
chengduoZH 已提交
251 252 253 254 255 256
the feature, and W is the width of the feature.
Filters(Input) is MCDHW format, where M is the number of output image channels,
C is the number of input image channels, D is the depth of the filter,
H is the height of the filter, and W is the width of the filter.
Parameters(strides, paddings, dilations) are three elements. These three elements
represent depth, height and width, respectively.
C
fix doc  
chengduoZH 已提交
257 258 259 260
The input(X) size and output(Out) size may be different.

Example:
  Input:
C
chengduoZH 已提交
261 262
       Input shape: $(N, C_{in}, D_{in}, H_{in}, W_{in})$
       Filter shape: $(C_{out}, C_{in}, D_f, H_f, W_f)$
C
fix doc  
chengduoZH 已提交
263
  Output:
C
chengduoZH 已提交
264 265 266 267 268 269 270
       Output shape: $(N, C_{out}, D_{out}, H_{out}, W_{out})$
  Where
  $$
       D_{out}= \frac{(D_{in} + 2 * paddings[0] - (dilations[0] * (D_f - 1) + 1))}{ strides[0]}+ 1 \\
       H_{out}= \frac{(H_{in} + 2 * paddings[1] - (dilations[1] * (H_f - 1) + 1))}{ strides[1]}+ 1 \\
       W_{out}= \frac{(W_{in} + 2 * paddings[2] - (dilations[2] * (W_f - 1) + 1))}{ strides[2]}+ 1
  $$
C
chengduoZH 已提交
271 272 273
)DOC");
}

C
chengduoZH 已提交
274 275 276 277 278 279 280 281 282 283 284
void ConvOpGrad::InferShape(framework::InferShapeContext* ctx) const {
  auto in_dims = ctx->GetInputDim("Input");
  auto filter_dims = ctx->GetInputDim("Filter");
  if (ctx->HasOutput(framework::GradVarName("Input"))) {
    ctx->SetOutputDim(framework::GradVarName("Input"), in_dims);
  }
  if (ctx->HasOutput(framework::GradVarName("Filter"))) {
    ctx->SetOutputDim(framework::GradVarName("Filter"), filter_dims);
  }
}

285 286 287
framework::OpKernelType ConvOpGrad::GetExpectedKernelType(
    const framework::ExecutionContext& ctx) const {
  bool use_cudnn = ctx.Attr<bool>("use_cudnn");
288
  use_cudnn &= platform::is_gpu_place(ctx.GetPlace());
C
chengduoZH 已提交
289 290 291 292 293 294 295
#ifdef PADDLE_WITH_CUDA
  if (platform::is_gpu_place(ctx.GetPlace())) {
    auto& dev_ctx = ctx.template device_context<platform::CUDADeviceContext>();
    use_cudnn &= dev_ctx.cudnn_handle() != nullptr;
  }
#endif

296 297 298 299 300 301 302 303 304 305 306 307 308 309
  framework::LibraryType library_;
  if (use_cudnn) {
    library_ = framework::LibraryType::kCUDNN;
  } else {
    library_ = framework::LibraryType::kPlain;
  }

  std::string data_format = ctx.Attr<std::string>("data_format");
  framework::DataLayout layout_ = framework::StringToDataLayout(data_format);
  return framework::OpKernelType(
      framework::ToDataType(ctx.Input<Tensor>("Input")->type()), ctx.GetPlace(),
      layout_, library_);
}

C
chengduoZH 已提交
310 311 312 313
}  // namespace operators
}  // namespace paddle

namespace ops = paddle::operators;
C
chengduoZH 已提交
314 315
REGISTER_OP(conv2d, ops::ConvOp, ops::Conv2DOpMaker, conv2d_grad,
            ops::ConvOpGrad);
316 317

// depthwise convolution op
318 319
REGISTER_OP(depthwise_conv2d, ops::ConvOp, ops::Conv2DOpMaker,
            depthwise_conv2d_grad, ops::ConvOpGrad);
C
chengduoZH 已提交
320 321 322
REGISTER_OP(conv3d, ops::ConvOp, ops::Conv3DOpMaker, conv3d_grad,
            ops::ConvOpGrad);

323 324
// depthwise conv kernel
// TODO(xingzhaolong): neon kernel for mobile
Z
zlx 已提交
325
REGISTER_OP_CPU_KERNEL(
326
    depthwise_conv2d,
X
xzl 已提交
327 328 329 330
    ops::GemmConvKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvKernel<paddle::platform::CPUDeviceContext, double>);

REGISTER_OP_CPU_KERNEL(
331
    depthwise_conv2d_grad,
X
xzl 已提交
332 333
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, double>);
Z
zlx 已提交
334

C
chengduoZH 已提交
335
REGISTER_OP_CPU_KERNEL(
Q
QI JUN 已提交
336 337 338 339 340 341
    conv2d, ops::GemmConvKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvKernel<paddle::platform::CPUDeviceContext, double>);
REGISTER_OP_CPU_KERNEL(
    conv2d_grad,
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, double>);
C
chengduoZH 已提交
342 343

REGISTER_OP_CPU_KERNEL(
Q
QI JUN 已提交
344 345 346 347 348 349
    conv3d, ops::GemmConvKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvKernel<paddle::platform::CPUDeviceContext, double>);
REGISTER_OP_CPU_KERNEL(
    conv3d_grad,
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvGradKernel<paddle::platform::CPUDeviceContext, double>);