Add GroupNorm forward operation (#2623)

Author: kyeonghwanryu

docs/apireference.rst

@@ -24,4 +24,5 @@ API Reference
     layernorm
     sum
     argmax
+    groupnorm
 

docs/groupnorm.rst

@@ -0,0 +1,20 @@
+
+GroupNorm Layer(experimental)
+=============================
+
+The groupnorm types and functions.
+It splits input channels into num_group groups and do normalize for each group.
+
+To enable this, define MIOPEN_BETA_API before including miopen.h.
+
+
+miopenNormMode_t
+-----------------------
+
+.. doxygenenum::  miopenNormMode_t
+
+miopenGroupNormForward
+----------------------------------
+
+.. doxygenfunction::  miopenGroupNormForward
+

docs/layernorm.rst

@@ -6,10 +6,10 @@ The layernorm  types and functions.
 To enable this, define MIOPEN_BETA_API before including miopen.h.
 
 
-miopenLayerNormMode_t
+miopenNormMode_t
 -----------------------
 
-.. doxygenenum::  miopenLayerNormMode_t
+.. doxygenenum::  miopenNormMode_t
 
 miopenLayerNormForward
 ----------------------------------

driver/driver.hpp

@@ -151,7 +151,7 @@ inline void PadBufferSize(size_t& sz, int datatype_sz)
            "pool[fp16], lrn[fp16], "
            "activ[fp16], softmax[fp16], bnorm[fp16], rnn[fp16], gemm[fp16], ctc, dropout[fp16], "
            "tensorop[fp16], reduce[fp16|fp64], layernorm[bfp16|fp16], sum[bfp16|fp16], "
-           "argmax[bfp16|fp16]\n");
+           "argmax[bfp16|fp16], groupnorm[bfp16|fp16]\n");
     exit(0); // NOLINT (concurrency-mt-unsafe)
 }
 
@@ -175,6 +175,7 @@ inline std::string ParseBaseArg(int argc, char* argv[])
        arg != "reduce" && arg != "reducefp16" && arg != "reducefp64" && arg != "layernorm" &&
        arg != "layernormfp16" && arg != "layernormbfp16" && arg != "sum" && arg != "sumfp16" &&
        arg != "sumbfp16" && arg != "argmax" && arg != "argmaxfp16" && arg != "argmaxbfp16" &&
+       arg != "groupnorm" && arg != "groupnormfp16" && arg != "groupnormbfp16" &&
        arg != "--version")
     {
         printf("FAILED: Invalid Base Input Argument\n");