From 8740019e0e167a9d63b0546c5c9de64f15903c7e Mon Sep 17 00:00:00 2001 From: Sandeep Krishnamurthy Date: Thu, 24 Jan 2019 15:44:22 -0800 Subject: [PATCH] Move back changes to original image operators files --- src/operator/image/image_random-inl.h | 102 ++++++++++++++++++ src/operator/image/image_random.cc | 39 +++++++ src/operator/image/image_random.cu | 44 ++++---- src/operator/image/totensor_op-inl.h | 142 -------------------------- src/operator/image/totensor_op.cc | 71 ------------- src/operator/image/totensor_op.cu | 30 ------ 6 files changed, 164 insertions(+), 264 deletions(-) delete mode 100644 src/operator/image/totensor_op-inl.h delete mode 100644 src/operator/image/totensor_op.cc delete mode 100644 src/operator/image/totensor_op.cu diff --git a/src/operator/image/image_random-inl.h b/src/operator/image/image_random-inl.h index ae5acb20ac86..9159ee431710 100644 --- a/src/operator/image/image_random-inl.h +++ b/src/operator/image/image_random-inl.h @@ -39,6 +39,108 @@ namespace mxnet { namespace op { namespace image { +// There are no parameters for this operator. +// Hence, no arameter registration. + +// Shape and Type inference for image to tensor operator +inline bool ToTensorShape(const nnvm::NodeAttrs& attrs, + std::vector *in_attrs, + std::vector *out_attrs) { + CHECK_EQ(in_attrs->size(), 1U); + CHECK_EQ(out_attrs->size(), 1U); + + TShape &shp = (*in_attrs)[0]; + if (!shp.ndim()) return false; + + CHECK((shp.ndim() == 3) || (shp.ndim() == 4)) + << "Input image must have shape (height, width, channels), or " + << "(N, height, width, channels) but got " << shp; + if (shp.ndim() == 3) { + SHAPE_ASSIGN_CHECK(*out_attrs, 0, TShape({shp[2], shp[0], shp[1]})); + } else if (shp.ndim() == 4) { + SHAPE_ASSIGN_CHECK(*out_attrs, 0, TShape({shp[0], shp[3], shp[1], shp[2]})); + } + + return true; +} + +inline bool ToTensorType(const nnvm::NodeAttrs& attrs, + std::vector *in_attrs, + std::vector *out_attrs) { + CHECK_EQ(in_attrs->size(), 1U); + CHECK_EQ(out_attrs->size(), 1U); + TYPE_ASSIGN_CHECK(*out_attrs, 0, mshadow::kFloat32); + return (*in_attrs)[0] != -1; +} + +// Operator Implementation + +template +struct totensor_forward { + template + MSHADOW_XINLINE static void Map(int l, float* out_data, const DType* in_data, + const int c, const int length, const int channel, + const int step, const float normalize_factor = 255.0f) { + KERNEL_ASSIGN(out_data[step + c*length + l], req, + (in_data[step + l*channel + c]) / normalize_factor); + } +}; + +template +void ToTensorImpl(const OpContext &ctx, + const std::vector &inputs, + const std::vector &outputs, + const std::vector &req, + const int length, + const int channel, + const int step = 0) { + mshadow::Stream *s = ctx.get_stream(); + + MSHADOW_TYPE_SWITCH(inputs[0].type_flag_, DType, { + MXNET_ASSIGN_REQ_SWITCH(req[0], req_type, { + float* output = outputs[0].dptr(); + DType* input = inputs[0].dptr(); + + for (int c = 0; c < channel; ++c) { + mxnet_op::Kernel, xpu>::Launch( + s, length, output, input, c, length, channel, step); + } + }); + }); +} + +template +void ToTensorOpForward(const nnvm::NodeAttrs &attrs, + const OpContext &ctx, + const std::vector &inputs, + const std::vector &req, + const std::vector &outputs) { + CHECK_EQ(inputs.size(), 1U); + CHECK_EQ(outputs.size(), 1U); + CHECK_EQ(req.size(), 1U); + + CHECK_EQ(req[0], kWriteTo) + << "`to_tensor` does not support inplace updates"; + + // 3D Input - (h, w, c) + if (inputs[0].ndim() == 3) { + const int length = inputs[0].shape_[0] * inputs[0].shape_[1]; + const int channel = inputs[0].shape_[2]; + ToTensorImpl(ctx, inputs, outputs, req, length, channel); + } else if (inputs[0].ndim() == 4) { + // 4D input (n, h, w, c) + const int batch_size = inputs[0].shape_[0]; + const int length = inputs[0].shape_[1] * inputs[0].shape_[2]; + const int channel = inputs[0].shape_[3]; + const int step = channel * length; + + #pragma omp parallel for + for (auto n = 0; n < batch_size; ++n) { + ToTensorImpl(ctx, inputs, outputs, req, length, channel, n*step); + } + } +} + struct NormalizeParam : public dmlc::Parameter { nnvm::Tuple mean; nnvm::Tuple std; diff --git a/src/operator/image/image_random.cc b/src/operator/image/image_random.cc index 56df9e37a3bb..f68a43aa8af5 100644 --- a/src/operator/image/image_random.cc +++ b/src/operator/image/image_random.cc @@ -38,6 +38,45 @@ DMLC_REGISTER_PARAMETER(AdjustLightingParam); DMLC_REGISTER_PARAMETER(RandomLightingParam); DMLC_REGISTER_PARAMETER(RandomColorJitterParam); +NNVM_REGISTER_OP(_image_to_tensor) +.describe(R"code(Converts an image NDArray of shape (H x W x C) or (N x H x W x C) +with values in the range [0, 255] to a tensor NDArray of shape (C x H x W) or (N x C x H x W) +with values in the range [0, 1) + +Example: + .. code-block:: python + image = mx.nd.random.uniform(0, 255, (4, 2, 3)).astype(dtype=np.uint8) + to_tensor(image) + [[[ 0.85490197 0.72156864] + [ 0.09019608 0.74117649] + [ 0.61960787 0.92941177] + [ 0.96470588 0.1882353 ]] + [[ 0.6156863 0.73725492] + [ 0.46666667 0.98039216] + [ 0.44705883 0.45490196] + [ 0.01960784 0.8509804 ]] + [[ 0.39607844 0.03137255] + [ 0.72156864 0.52941179] + [ 0.16470589 0.7647059 ] + [ 0.05490196 0.70588237]]] + +)code" ADD_FILELINE) +.set_num_inputs(1) +.set_num_outputs(1) +.set_attr("FListInputNames", + [](const NodeAttrs& attrs) { + return std::vector{"data"}; + }) +.set_attr("FInferShape", ToTensorShape) +.set_attr("FInferType", ToTensorType) +.set_attr("FCompute", ToTensorOpForward) +.set_attr("FInplaceOption", + [](const NodeAttrs& attrs) { + return std::vector >{{0, 0}}; + }) +.set_attr("FGradient", ElemwiseGradUseNone{ "_copy" }) +.add_argument("data", "NDArray-or-Symbol", "Input ndarray"); + NNVM_REGISTER_OP(_image_normalize) .describe(R"code(Normalize an tensor of shape (C x H x W) or (N x C x H x W) with mean and standard deviation. diff --git a/src/operator/image/image_random.cu b/src/operator/image/image_random.cu index 404c3d25477a..5f9aff27e85b 100644 --- a/src/operator/image/image_random.cu +++ b/src/operator/image/image_random.cu @@ -1,26 +1,26 @@ /* - * Licensed to the Apache Software Foundation (ASF) under one - * or more contributor license agreements. See the NOTICE file - * distributed with this work for additional information - * regarding copyright ownership. The ASF licenses this file - * to you 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. - */ +* Licensed to the Apache Software Foundation (ASF) under one +* or more contributor license agreements. See the NOTICE file +* distributed with this work for additional information +* regarding copyright ownership. The ASF licenses this file +* to you 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. +*/ /*! - * \file image_random.cu - * \brief GPU Implementation of image transformation operators - */ +* \file image_random.cu +* \brief GPU Implementation of image transformation operators +*/ #include "./image_random-inl.h" #include "../elemwise_op_common.h" @@ -28,13 +28,15 @@ namespace mxnet { namespace op { namespace image { +NNVM_REGISTER_OP(_image_to_tensor) +.set_attr("FCompute", ToTensorOpForward); + NNVM_REGISTER_OP(_image_normalize) .set_attr("FCompute", NormalizeOpForward); NNVM_REGISTER_OP(_backward_image_normalize) .set_attr("FCompute", NormalizeOpBackward); - } // namespace image } // namespace op } // namespace mxnet diff --git a/src/operator/image/totensor_op-inl.h b/src/operator/image/totensor_op-inl.h deleted file mode 100644 index 14d7e0bf0126..000000000000 --- a/src/operator/image/totensor_op-inl.h +++ /dev/null @@ -1,142 +0,0 @@ -/* -* Licensed to the Apache Software Foundation (ASF) under one -* or more contributor license agreements. See the NOTICE file -* distributed with this work for additional information -* regarding copyright ownership. The ASF licenses this file -* to you 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. -*/ - -/*! - * Copyright (c) 2018 by Contributors - * \file totensor_op-inl.h - * \brief Image to tensor operator -*/ -#ifndef MXNET_OPERATOR_IMAGE_TOTENSOR_OP_INL_H_ -#define MXNET_OPERATOR_IMAGE_TOTENSOR_OP_INL_H_ - - -#include -#include "../elemwise_op_common.h" - -namespace mxnet { -namespace op { -namespace image { - -// There are no parameters for this operator. -// Hence, no arameter registration. - -// Shape and Type inference for image to tensor operator -inline bool ToTensorShape(const nnvm::NodeAttrs& attrs, - std::vector *in_attrs, - std::vector *out_attrs) { - CHECK_EQ(in_attrs->size(), 1U); - CHECK_EQ(out_attrs->size(), 1U); - - TShape &shp = (*in_attrs)[0]; - if (!shp.ndim()) return false; - - CHECK((shp.ndim() == 3) || (shp.ndim() == 4)) - << "Input image must have shape (height, width, channels), or " - << "(N, height, width, channels) but got " << shp; - if (shp.ndim() == 3) { - SHAPE_ASSIGN_CHECK(*out_attrs, 0, TShape({shp[2], shp[0], shp[1]})); - } else if (shp.ndim() == 4) { - SHAPE_ASSIGN_CHECK(*out_attrs, 0, TShape({shp[0], shp[3], shp[1], shp[2]})); - } - - return true; -} - -inline bool ToTensorType(const nnvm::NodeAttrs& attrs, - std::vector *in_attrs, - std::vector *out_attrs) { - CHECK_EQ(in_attrs->size(), 1U); - CHECK_EQ(out_attrs->size(), 1U); - TYPE_ASSIGN_CHECK(*out_attrs, 0, mshadow::kFloat32); - return (*in_attrs)[0] != -1; -} - -// Operator Implementation - -template -struct totensor_forward { - template - MSHADOW_XINLINE static void Map(int l, float* out_data, const DType* in_data, - const int c, const int length, const int channel, - const int step, const float normalize_factor = 255.0f) { - KERNEL_ASSIGN(out_data[step + c*length + l], req, - (in_data[step + l*channel + c]) / normalize_factor); - } -}; - -template -void ToTensorImpl(const OpContext &ctx, - const std::vector &inputs, - const std::vector &outputs, - const std::vector &req, - const int length, - const int channel, - const int step = 0) { - mshadow::Stream *s = ctx.get_stream(); - - MSHADOW_TYPE_SWITCH(inputs[0].type_flag_, DType, { - MXNET_ASSIGN_REQ_SWITCH(req[0], req_type, { - float* output = outputs[0].dptr(); - DType* input = inputs[0].dptr(); - - for (int c = 0; c < channel; ++c) { - mxnet_op::Kernel, xpu>::Launch( - s, length, output, input, c, length, channel, step); - } - }); - }); -} - -template -void ToTensorOpForward(const nnvm::NodeAttrs &attrs, - const OpContext &ctx, - const std::vector &inputs, - const std::vector &req, - const std::vector &outputs) { - CHECK_EQ(inputs.size(), 1U); - CHECK_EQ(outputs.size(), 1U); - CHECK_EQ(req.size(), 1U); - - CHECK_EQ(req[0], kWriteTo) - << "`to_tensor` does not support inplace updates"; - - // 3D Input - (h, w, c) - if (inputs[0].ndim() == 3) { - const int length = inputs[0].shape_[0] * inputs[0].shape_[1]; - const int channel = inputs[0].shape_[2]; - ToTensorImpl(ctx, inputs, outputs, req, length, channel); - } else if (inputs[0].ndim() == 4) { - // 4D input (n, h, w, c) - const int batch_size = inputs[0].shape_[0]; - const int length = inputs[0].shape_[1] * inputs[0].shape_[2]; - const int channel = inputs[0].shape_[3]; - const int step = channel * length; - - #pragma omp parallel for - for (auto n = 0; n < batch_size; ++n) { - ToTensorImpl(ctx, inputs, outputs, req, length, channel, n*step); - } - } -} - -} // namespace image -} // namespace op -} // namespace mxnet - -#endif // MXNET_OPERATOR_IMAGE_TOTENSOR_OP_INL_H_ diff --git a/src/operator/image/totensor_op.cc b/src/operator/image/totensor_op.cc deleted file mode 100644 index 1369a6815d71..000000000000 --- a/src/operator/image/totensor_op.cc +++ /dev/null @@ -1,71 +0,0 @@ -/* -* Licensed to the Apache Software Foundation (ASF) under one -* or more contributor license agreements. See the NOTICE file -* distributed with this work for additional information -* regarding copyright ownership. The ASF licenses this file -* to you 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. -*/ - -/*! - * \file totensor_op.cc - * \brief CPU Implementation of ToTensor op - */ -#include "./totensor_op-inl.h" - -namespace mxnet { -namespace op { -namespace image { - -NNVM_REGISTER_OP(_image_to_tensor) -.describe(R"code(Converts an image NDArray of shape (H x W x C) or (N x H x W x C) -with values in the range [0, 255] to a tensor NDArray of shape (C x H x W) or (N x C x H x W) -with values in the range [0, 1) - -Example: - .. code-block:: python - image = mx.nd.random.uniform(0, 255, (4, 2, 3)).astype(dtype=np.uint8) - to_tensor(image) - [[[ 0.85490197 0.72156864] - [ 0.09019608 0.74117649] - [ 0.61960787 0.92941177] - [ 0.96470588 0.1882353 ]] - [[ 0.6156863 0.73725492] - [ 0.46666667 0.98039216] - [ 0.44705883 0.45490196] - [ 0.01960784 0.8509804 ]] - [[ 0.39607844 0.03137255] - [ 0.72156864 0.52941179] - [ 0.16470589 0.7647059 ] - [ 0.05490196 0.70588237]]] - -)code" ADD_FILELINE) -.set_num_inputs(1) -.set_num_outputs(1) -.set_attr("FListInputNames", - [](const NodeAttrs& attrs) { - return std::vector{"data"}; - }) -.set_attr("FInferShape", ToTensorShape) -.set_attr("FInferType", ToTensorType) -.set_attr("FCompute", ToTensorOpForward) -.set_attr("FInplaceOption", - [](const NodeAttrs& attrs) { - return std::vector >{{0, 0}}; - }) -.set_attr("FGradient", ElemwiseGradUseNone{ "_copy" }) -.add_argument("data", "NDArray-or-Symbol", "Input ndarray"); - -} // namespace image -} // namespace op -} // namespace mxnet diff --git a/src/operator/image/totensor_op.cu b/src/operator/image/totensor_op.cu deleted file mode 100644 index 55084ac0d9c5..000000000000 --- a/src/operator/image/totensor_op.cu +++ /dev/null @@ -1,30 +0,0 @@ -/* -* Licensed to the Apache Software Foundation (ASF) under one -* or more contributor license agreements. See the NOTICE file -* distributed with this work for additional information -* regarding copyright ownership. The ASF licenses this file -* to you 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. -*/ -#include "./totensor_op-inl.h" - -namespace mxnet { -namespace op { -namespace image { - -NNVM_REGISTER_OP(_image_to_tensor) -.set_attr("FCompute", ToTensorOpForward); - -} // namespace image -} // namespace op -} // namespace mxnet