conv_transpose_op.cc 14.2 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"
S
Siddharth Goyal 已提交
16 17
#include <string>
#include <vector>
C
chengduoZH 已提交
18 19 20 21

namespace paddle {
namespace operators {

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

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

C
chengduoZH 已提交
36 37 38 39 40 41 42 43 44
  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 已提交
45
                    "ConvTransposeOp paddings dimension and strides "
C
chengduoZH 已提交
46
                    "dimension should be the same.");
C
chengduoZH 已提交
47 48 49
  PADDLE_ENFORCE_EQ(paddings.size(), dilations.size(),
                    "ConvTransposeOp paddings dimension and dilations "
                    "dimension should be the same.");
C
chengduoZH 已提交
50 51 52
  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 已提交
53

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

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

Y
Yu Yang 已提交
87
void Conv2DTransposeOpMaker::Make() {
C
chengduoZH 已提交
88 89 90 91
  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");
}

Y
Yu Yang 已提交
169
void Conv3DTransposeOpMaker::Make() {
C
chengduoZH 已提交
170 171 172 173 174 175
  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 已提交
176 177
  AddInput("Filter",
           "(Tensor) The filter tensor of convolution transpose operator."
C
chengduoZH 已提交
178 179 180
           "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 已提交
181 182
           "is the depth of the filter, H is the height of the filter, and "
           "W is the width of the filter."
C
chengduoZH 已提交
183
           "We enforce groups number == 1 and padding == 0 in "
C
chengduoZH 已提交
184
           "the convolution3d transpose scenario.");
C
chengduoZH 已提交
185 186 187 188
  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 已提交
189 190
            "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 已提交
191 192 193 194 195 196 197

  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 已提交
198
  AddAttr<std::vector<int>>("strides",
C
chengduoZH 已提交
199
                            "(vector<int> default:{1, 1, 1}), the "
200
                            "strides{d_stride, h_stride, w_stride} of "
C
chengduoZH 已提交
201
                            "convolution transpose operator.")
C
chengduoZH 已提交
202
      .SetDefault({1, 1, 1});
C
chengduoZH 已提交
203
  AddAttr<std::vector<int>>("paddings",
C
chengduoZH 已提交
204
                            "(vector<int> default:{0, 0, 0}), paddings(d_pad, "
C
chengduoZH 已提交
205
                            "h_pad, w_pad) of convolution transpose operator.")
C
chengduoZH 已提交
206
      .SetDefault({0, 0, 0});
207 208 209 210 211 212 213 214 215 216 217 218 219 220 221 222 223 224 225
  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 已提交
226
  AddComment(R"DOC(
C
chengduoZH 已提交
227 228
Convolution3D Transpose Operator.

C
chengduoZH 已提交
229
The convolution transpose operation calculates the output based on the input, filter
C
chengduoZH 已提交
230
and dilations, strides, paddings, groups parameters. The size of each dimension of the
C
chengduoZH 已提交
231
parameters is checked in the infer-shape.
C
chengduoZH 已提交
232 233 234 235 236 237 238 239
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 已提交
240
The input(X) size and output(Out) size may be different.
C
chengduoZH 已提交
241 242

Example:   
C
chengduoZH 已提交
243
  Input:
C
chengduoZH 已提交
244 245
       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 已提交
246
  Output:
C
chengduoZH 已提交
247 248 249
       Output shape: $(N, C_{out}, D_{out}, H_{out}, W_{out})$
  Where
  $$
250 251 252
       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 已提交
253
  $$
C
chengduoZH 已提交
254 255 256
)DOC");
}

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

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

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

Y
Yang Yang 已提交
297 298
REGISTER_OPERATOR(conv2d_transpose, ops::ConvTransposeOp,
                  ops::Conv2DTransposeOpMaker,
299 300
                  paddle::framework::DefaultGradOpDescMaker<true>);
REGISTER_OPERATOR(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

Y
Yang Yang 已提交
312 313
REGISTER_OPERATOR(conv3d_transpose, ops::ConvTransposeOp,
                  ops::Conv3DTransposeOpMaker,
314 315
                  paddle::framework::DefaultGradOpDescMaker<true>);
REGISTER_OPERATOR(conv3d_transpose_grad, ops::ConvTransposeOpGrad);
C
chengduoZH 已提交
316 317

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