From 9b9f33d37e380a26427c5d7d257543ce5554ba93 Mon Sep 17 00:00:00 2001
From: Khushbu Agarwal <khuagarw@amd.com>
Date: Thu, 3 Apr 2025 16:24:34 -0700
Subject: [PATCH] Documentation for newly added struct (#2051)

[ROCm/composable_kernel commit: b443056a26cd25e6e621ff1c026b02eefdfe1f29]
---
 tile_engine/ops/gemm/gemm_host_api.hpp        | 17 ++++++++++++++++-
 tile_engine/ops/gemm/gemm_instance_builder.py |  3 +--
 2 files changed, 17 insertions(+), 3 deletions(-)

diff --git a/tile_engine/ops/gemm/gemm_host_api.hpp b/tile_engine/ops/gemm/gemm_host_api.hpp
index 4f0ea52a18..3fa6dca863 100644
--- a/tile_engine/ops/gemm/gemm_host_api.hpp
+++ b/tile_engine/ops/gemm/gemm_host_api.hpp
@@ -54,6 +54,17 @@ struct DataTypeTraits<ck_tile::pk_int4_t>
     static constexpr const char* name = "pk_int4_t";
 };
 
+/**
+ * @brief  trait for GEMM kernel
+    * @param pipeline:   pipeline name
+    * @param scheduler:  scheduler name
+    * @param epilogue:  epilogue name
+    * @param kPadM:     padding for M dimension
+    * @param kPadN:     padding for N dimension
+    * @param kPadK:     padding for K dimension 
+ * 
+ */
+
 struct KernelTraits
 {
     std::string pipeline;
@@ -173,7 +184,11 @@ void permute_vectors_i4x4_b(Tensor& tensor)
     }
 }
 
-// verification code
+/**
+ * @brief Function to verify the kernel output with reference implementation on CPU/GPU
+ * 
+ */
+
 template <typename ADataType,
           typename BDataType,
           typename AccDataType,
diff --git a/tile_engine/ops/gemm/gemm_instance_builder.py b/tile_engine/ops/gemm/gemm_instance_builder.py
index c0dad03ef0..e449dff94d 100755
--- a/tile_engine/ops/gemm/gemm_instance_builder.py
+++ b/tile_engine/ops/gemm/gemm_instance_builder.py
@@ -555,11 +555,10 @@ def do_gen_blobs(args, gemm_config):
      
 
 def main(args):
-    # Read and validate json file
+    # Read json file
     with open(args.json, 'r') as json_file:
         config_data = json.load(json_file)
     
-    # Validate and parse configuration
     gemm_config = GemmConfig(config_data)
 
     if args.list_blobs: