new add l2normalize_grad gpu ops.

mindspore/ccsrc/backend/kernel_compiler/gpu/nn/l2normalize_gpu_kernel.h

@@ -159,19 +159,19 @@ class L2NormalizeGpuKernel : public GpuKernel {
     CHECK_CUDNN_RET_WITH_EXCEPT(cudnnGetTensorSizeInBytes(inputA_descriptor_, &input_size_),
                                 "cudnnGetTensorSizeInBytes failed.");
     input_size_list_.push_back(input_size_);
-    MS_LOG(ERROR) << "input size: " << input_size_;
+
     output_size_list_.push_back(output_size_);
-    MS_LOG(ERROR) << "output size: " << output_size_;
+
     CHECK_CUDNN_RET_WITH_EXCEPT(cudnnGetTensorSizeInBytes(outputC_descriptor_, &workspace_size_),
                                 "cudnnGetTensorSizeInBytes failed.");
     workspace_size_list_.push_back(workspace_size_);
-    MS_LOG(ERROR) << "workspace_size_1 size: " << workspace_size_;
+
     CHECK_CUDNN_RET_WITH_EXCEPT(
       cudnnGetReductionWorkspaceSize(cudnn_handle_, reduce_tensor_descriptor_, inputA_descriptor_, outputC_descriptor_,
                                      &workspace_size_),
       "cudnnGetReductionWorkspaceSize failed.");
     workspace_size_list_.push_back(workspace_size_);
-    MS_LOG(ERROR) << "workspace_size_2 size: " << workspace_size_;
+
     return;
   }
 

mindspore/ccsrc/backend/kernel_compiler/gpu/nn/l2normalize_grad_gpu_kernel.cc

@@ -0,0 +1,43 @@
+/**
+ * Copyright 2020 Huawei Technologies Co., Ltd
+ *
+ * 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
+ *
+ * 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 "backend/kernel_compiler/gpu/nn/l2normalize_grad_gpu_kernel.h"
+
+namespace mindspore {
+namespace kernel {
+MS_REG_GPU_KERNEL_ONE(L2NormalizeGrad,
+                      KernelAttr()
+                        .AddInputAttr(kNumberTypeFloat32)
+                        .AddInputAttr(kNumberTypeFloat32)
+                        .AddInputAttr(kNumberTypeFloat32)
+                        .AddOutputAttr(kNumberTypeFloat32),
+                      L2NormalizeGradGpuKernel, float)
+MS_REG_GPU_KERNEL_ONE(L2NormalizeGrad,
+                      KernelAttr()
+                        .AddInputAttr(kNumberTypeFloat16)
+                        .AddInputAttr(kNumberTypeFloat16)
+                        .AddInputAttr(kNumberTypeFloat16)
+                        .AddOutputAttr(kNumberTypeFloat16),
+                      L2NormalizeGradGpuKernel, half)
+MS_REG_GPU_KERNEL_ONE(L2NormalizeGrad,
+                      KernelAttr()
+                        .AddInputAttr(kNumberTypeInt32)
+                        .AddInputAttr(kNumberTypeInt32)
+                        .AddInputAttr(kNumberTypeInt32)
+                        .AddOutputAttr(kNumberTypeInt32),
+                      L2NormalizeGradGpuKernel, int)
+}  // namespace kernel
+}  // namespace mindspore

tests/st/ops/gpu/test_l2normalize_grad_op.py

@@ -0,0 +1,52 @@
+# Copyright 2020 Huawei Technologies Co., Ltd
+#
+# 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
+#
+# 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.
+# ============================================================================
+
+import numpy as np
+import pytest
+
+import mindspore.context as context
+from mindspore.common.tensor import Tensor
+from mindspore.nn import Cell
+from mindspore.ops.operations import _grad_ops as G
+
+class Net(Cell):
+    def __init__(self, axis=0, epsilon=1e-12):
+        super(Net, self).__init__()
+        self.norm_grad = G.L2NormalizeGrad(axis=axis, epsilon=epsilon)
+
+    def construct(self, x, out, dout):
+        return self.norm_grad(x, out, dout)
+
+
+@pytest.mark.level0
+@pytest.mark.platform_x86_gpu_training
+@pytest.mark.env_onecard
+def test_l2normalize_grad():
+    axis_ = 0
+    x = np.random.randint(1, 10, (2, 3, 4, 4)).astype(np.float32)
+    y = x / np.sqrt(np.sum(x**2, axis=axis_, keepdims=True))
+    dy = np.random.randint(1, 10, (2, 3, 4, 4)).astype(np.float32)
+    expect = (dy - y * np.sum(y * dy, axis=axis_, keepdims=True)) / np.sqrt(np.sum(x**2, axis=axis_, keepdims=True))
+    x = Tensor(x)
+    y = Tensor(y)
+    dy = Tensor(dy)
+    error = np.ones(shape=[2, 3, 4, 4]) * 1.0e-5
+
+    context.set_context(mode=context.GRAPH_MODE, device_target="GPU")
+    norm_grad_op = Net(axis=axis_)
+    output = norm_grad_op(x, y, dy)
+    diff = output.asnumpy() - expect
+    assert np.all(diff < error)
+    assert np.all(-diff < error)