conv_transpose_op.cc 14.3 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_transpose_op.h"
C
chengduoZH 已提交
16 17 18 19

namespace paddle {
namespace operators {

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

  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");
C
chengduoZH 已提交
32
  std::vector<int> dilations = ctx->Attrs().Get<std::vector<int>>("dilations");
C
chengduoZH 已提交
33

C
chengduoZH 已提交
34 35 36 37 38 39 40 41 42
  PADDLE_ENFORCE(in_dims.size() == 4 || in_dims.size() == 5,
                 "ConvTransposeOp intput should be 4-D or 5-D tensor.");
  PADDLE_ENFORCE_EQ(in_dims.size(), filter_dims.size(),
                    "ConvTransposeOp input dimension and filter dimension "
                    "should be the same.");
  PADDLE_ENFORCE(in_dims.size() - strides.size() == 2U,
                 "ConvTransposeOp input dimension and strides dimension should "
                 "be consistent.");
  PADDLE_ENFORCE_EQ(paddings.size(), strides.size(),
C
chengduoZH 已提交
43
                    "ConvTransposeOp paddings dimension and strides "
C
chengduoZH 已提交
44
                    "dimension should be the same.");
C
chengduoZH 已提交
45 46 47
  PADDLE_ENFORCE_EQ(paddings.size(), dilations.size(),
                    "ConvTransposeOp paddings dimension and dilations "
                    "dimension should be the same.");
C
chengduoZH 已提交
48 49 50
  PADDLE_ENFORCE_EQ(in_dims[1], filter_dims[0],
                    "In ConvTransposeOp, The input channel should be the same "
                    "as the number of filters.");
C
chengduoZH 已提交
51

C
chengduoZH 已提交
52
  std::vector<int64_t> output_shape({in_dims[0], filter_dims[1]});
C
chengduoZH 已提交
53
  for (size_t i = 0; i < strides.size(); ++i) {
C
chengduoZH 已提交
54
    auto filter_extent = dilations[i] * (filter_dims[i + 2] - 1) + 1;
C
chengduoZH 已提交
55
    output_shape.push_back((in_dims[i + 2] - 1) * strides[i] - 2 * paddings[i] +
C
chengduoZH 已提交
56
                           filter_extent);
C
chengduoZH 已提交
57
  }
C
chengduoZH 已提交
58
  ctx->SetOutputDim("Output", framework::make_ddim(output_shape));
C
chengduoZH 已提交
59 60
}

61 62 63
framework::OpKernelType ConvTransposeOp::GetExpectedKernelType(
    const framework::ExecutionContext& ctx) const {
  bool use_cudnn = ctx.Attr<bool>("use_cudnn");
C
chengduoZH 已提交
64
  use_cudnn &= platform::is_gpu_place(ctx.GetPlace());
C
chengduoZH 已提交
65 66 67 68 69 70
#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
71 72 73 74 75 76 77 78 79 80 81 82 83 84
  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_);
}

85 86
Conv2DTransposeOpMaker::Conv2DTransposeOpMaker(OpProto* proto,
                                               OpAttrChecker* op_checker)
C
chengduoZH 已提交
87 88 89 90 91
    : OpProtoAndCheckerMaker(proto, op_checker) {
  AddInput(
      "Input",
      "(Tensor) The input tensor of convolution transpose operator. "
      "The format of input tensor is NCHW. Where N is batch size, C is the "
C
chengduoZH 已提交
92 93
      "number of input channels, H is the height of the feature, and "
      "W is the width of the feature.");
C
chengduoZH 已提交
94 95 96 97 98 99 100 101
  AddInput(
      "Filter",
      "(Tensor) The filter tensor of convolution transpose operator. "
      "The format of the filter tensor is MCHW, where M is the number of "
      "input feature channels, C is the number of "
      "output feature channels,"
      "H is the height of the filter, and W is the width of the filter. "
      "We enforce groups number == 1 in the convolution transpose scenario.");
C
chengduoZH 已提交
102
  AddOutput("Output",
C
chengduoZH 已提交
103
            "(Tensor) The output tensor of convolution transpose operator. "
C
chengduoZH 已提交
104
            "The format of output tensor is also NCHW.");
C
chengduoZH 已提交
105 106 107 108 109 110

  AddAttr<std::vector<int>>("dilations",
                            "(vector<int> default:{1, 1}), the "
                            "dilations(h_dilation, w_dilation) of convolution "
                            "transpose operator.")
      .SetDefault({1, 1});
C
chengduoZH 已提交
111 112
  AddAttr<std::vector<int>>(
      "strides",
C
chengduoZH 已提交
113
      "(vector<int> default:{1, 1}), the strides(h_stride, w_stride) of "
114
      "convolution transpose operator.")
C
chengduoZH 已提交
115
      .SetDefault({1, 1});
C
chengduoZH 已提交
116 117
  AddAttr<std::vector<int>>(
      "paddings",
C
chengduoZH 已提交
118
      "(vector<int> default:{0, 0}), the paddings(h_pad, w_pad) of convolution "
C
chengduoZH 已提交
119
      "transpose operator.")
C
chengduoZH 已提交
120
      .SetDefault({0, 0});
121 122 123 124 125 126 127 128 129 130 131 132 133 134 135 136 137 138 139
  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",
               "Used in cudnn kernel only. 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 hardward. This size should be carefully setted.")
      .SetDefault(4096);
C
chengduoZH 已提交
140
  AddComment(R"DOC(
C
chengduoZH 已提交
141 142
Convolution2D Transpose Operator.

C
chengduoZH 已提交
143
The convolution transpose operation calculates the output based on the input, filter
C
chengduoZH 已提交
144
and dilations, strides, paddings, groups parameters. The size of each dimension of the
C
chengduoZH 已提交
145
parameters is checked in the infer-shape.
C
chengduoZH 已提交
146 147 148 149 150 151 152
Input(Input) and output(Output) are in NCHW format. Where N is batchsize, C is the
number of channels, H is the height of the feature, and W is the width of the feature.
Filter(Input) is in MCHW format. Where M is the number of input feature channels,
C is the number of output feature channels, H is the height of the filter,
and W is the width of the filter.
Parameters(strides, paddings) are two elements. These two elements represent height
and width, respectively.
C
chengduoZH 已提交
153
The input(X) size and output(Out) size may be different.
C
chengduoZH 已提交
154

C
chengduoZH 已提交
155 156
Example:
  Input:
C
chengduoZH 已提交
157 158
       Input shape: $(N, C_{in}, H_{in}, W_{in})$
       Filter shape: $(C_{in}, C_{out}, H_f, W_f)$
C
chengduoZH 已提交
159
  Output:
C
chengduoZH 已提交
160 161 162
       Output shape: $(N, C_{out}, H_{out}, W_{out})$
  Where
  $$
163 164
       H_{out} = (H_{in} - 1) * strides[0] - 2 * paddings[0] + dilations[0] * (H_f - 1) + 1 \\
       W_{out} = (W_{in} - 1) * strides[1] - 2 * paddings[1] + dilations[1] * (W_f - 1) + 1
C
chengduoZH 已提交
165
  $$
C
chengduoZH 已提交
166 167 168
)DOC");
}

169 170
Conv3DTransposeOpMaker::Conv3DTransposeOpMaker(OpProto* proto,
                                               OpAttrChecker* op_checker)
C
chengduoZH 已提交
171
    : OpProtoAndCheckerMaker(proto, op_checker) {
C
chengduoZH 已提交
172 173 174 175 176 177
  AddInput("Input",
           "(Tensor) The input tensor of convolution transpose operator."
           "The format of input tensor is NCDHW. Where N is batch size, C is "
           "the 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 已提交
178 179
  AddInput("Filter",
           "(Tensor) The filter tensor of convolution transpose operator."
C
chengduoZH 已提交
180 181 182
           "The format of the filter tensor is MCDHW, where M is the number of "
           "input feature channels, C is the number of "
           "output feature channels, D "
C
chengduoZH 已提交
183 184
           "is the depth of the filter, H is the height of the filter, and "
           "W is the width of the filter."
C
chengduoZH 已提交
185
           "We enforce groups number == 1 and padding == 0 in "
C
chengduoZH 已提交
186
           "the convolution3d transpose scenario.");
C
chengduoZH 已提交
187 188 189 190
  AddOutput("Output",
            "(Tensor) The output tensor of convolution transpose operator."
            "The format of output tensor is also NCDHW."
            "Where N is batch size, C is "
C
chengduoZH 已提交
191 192
            "the 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 已提交
193 194 195 196 197 198 199

  AddAttr<std::vector<int>>(
      "dilations",
      "(vector<int> default:{1, 1, 1}), the "
      "dilations(d_dilation,h_dilation, w_dilation) of convolution "
      "transpose operator.")
      .SetDefault({1, 1, 1});
C
chengduoZH 已提交
200
  AddAttr<std::vector<int>>("strides",
C
chengduoZH 已提交
201
                            "(vector<int> default:{1, 1, 1}), the "
202
                            "strides{d_stride, h_stride, w_stride} of "
C
chengduoZH 已提交
203
                            "convolution transpose operator.")
C
chengduoZH 已提交
204
      .SetDefault({1, 1, 1});
C
chengduoZH 已提交
205
  AddAttr<std::vector<int>>("paddings",
C
chengduoZH 已提交
206
                            "(vector<int> default:{0, 0, 0}), paddings(d_pad, "
C
chengduoZH 已提交
207
                            "h_pad, w_pad) of convolution transpose operator.")
C
chengduoZH 已提交
208
      .SetDefault({0, 0, 0});
209 210 211 212 213 214 215 216 217 218 219 220 221 222 223 224 225 226 227
  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",
               "Used in cudnn kernel only. 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 hardward. This size should be carefully setted.")
      .SetDefault(4096);
C
chengduoZH 已提交
228
  AddComment(R"DOC(
C
chengduoZH 已提交
229 230
Convolution3D Transpose Operator.

C
chengduoZH 已提交
231
The convolution transpose operation calculates the output based on the input, filter
C
chengduoZH 已提交
232
and dilations, strides, paddings, groups parameters. The size of each dimension of the
C
chengduoZH 已提交
233
parameters is checked in the infer-shape.
C
chengduoZH 已提交
234 235 236 237 238 239 240 241
Input(Input) and output(Output) are in NCDHW format. Where N is batch size, C is the
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.
Filter(Input) is in MCDHW format. Where M is the number of input feature channels,
C is the number of output feature 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) are three elements. These three elements represent
depth, height and width, respectively.
C
chengduoZH 已提交
242
The input(X) size and output(Out) size may be different.
C
chengduoZH 已提交
243 244

Example:   
C
chengduoZH 已提交
245
  Input:
C
chengduoZH 已提交
246 247
       Input shape: $(N, C_{in}, D_{in}, H_{in}, W_{in})$
       Filter shape: $(C_{in}, C_{out}, D_f, H_f, W_f)$
C
chengduoZH 已提交
248
  Output:
C
chengduoZH 已提交
249 250 251
       Output shape: $(N, C_{out}, D_{out}, H_{out}, W_{out})$
  Where
  $$
252 253 254
       D_{out} = (D_{in} - 1) * strides[0] - 2 * paddings[0] + dilations[0] * (D_f - 1) + 1 \\
       H_{out} = (H_{in} - 1) * strides[1] - 2 * paddings[1] + dilations[1] * (H_f - 1) + 1 \\
       W_{out} = (W_{in} - 1) * strides[2] - 2 * paddings[2] + dilations[2] * (W_f - 1) + 1
C
chengduoZH 已提交
255
  $$
C
chengduoZH 已提交
256 257 258
)DOC");
}

C
chengduoZH 已提交
259
void ConvTransposeOpGrad::InferShape(framework::InferShapeContext* ctx) const {
C
chengduoZH 已提交
260 261 262 263 264 265 266 267 268 269
  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);
  }
}

270 271 272
framework::OpKernelType ConvTransposeOpGrad::GetExpectedKernelType(
    const framework::ExecutionContext& ctx) const {
  bool use_cudnn = ctx.Attr<bool>("use_cudnn");
273
  use_cudnn &= platform::is_gpu_place(ctx.GetPlace());
C
chengduoZH 已提交
274 275 276 277 278 279
#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
280 281 282 283 284 285 286 287 288 289 290 291 292 293
  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 已提交
294 295 296 297
}  // namespace operators
}  // namespace paddle

namespace ops = paddle::operators;
C
chengduoZH 已提交
298

C
chengduoZH 已提交
299 300
REGISTER_OP(conv2d_transpose, ops::ConvTransposeOp, ops::Conv2DTransposeOpMaker,
            conv2d_transpose_grad, ops::ConvTransposeOpGrad);
C
chengduoZH 已提交
301 302

REGISTER_OP_CPU_KERNEL(
C
chengduoZH 已提交
303
    conv2d_transpose,
Q
QI JUN 已提交
304 305
    ops::GemmConvTransposeKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvTransposeKernel<paddle::platform::CPUDeviceContext, double>);
C
chengduoZH 已提交
306
REGISTER_OP_CPU_KERNEL(
C
chengduoZH 已提交
307
    conv2d_transpose_grad,
Q
QI JUN 已提交
308 309 310
    ops::GemmConvTransposeGradKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvTransposeGradKernel<paddle::platform::CPUDeviceContext,
                                     double>);
C
chengduoZH 已提交
311

C
chengduoZH 已提交
312 313
REGISTER_OP(conv3d_transpose, ops::ConvTransposeOp, ops::Conv3DTransposeOpMaker,
            conv3d_transpose_grad, ops::ConvTransposeOpGrad);
C
chengduoZH 已提交
314 315

REGISTER_OP_CPU_KERNEL(
C
chengduoZH 已提交
316
    conv3d_transpose,
Q
QI JUN 已提交
317 318
    ops::GemmConvTransposeKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvTransposeKernel<paddle::platform::CPUDeviceContext, double>);
C
chengduoZH 已提交
319
REGISTER_OP_CPU_KERNEL(
C
chengduoZH 已提交
320
    conv3d_transpose_grad,
Q
QI JUN 已提交
321 322 323
    ops::GemmConvTransposeGradKernel<paddle::platform::CPUDeviceContext, float>,
    ops::GemmConvTransposeGradKernel<paddle::platform::CPUDeviceContext,
                                     double>);