Add script to generate CK profiler commands from the KTN data.

This commit is contained in:
Ville Pietilä
2025-06-13 13:49:38 +00:00
parent e314e6faf0
commit 3b8dfae5bc
3 changed files with 208 additions and 91 deletions

View File

@@ -224,7 +224,7 @@ def get_parser():
help="Input Layout (Default=NCHW for 2d conv, NCDHW for 3d conv)"
)
parser.add_argument(
"-forw",
"--forw",
"-F",
default=0,
type=int,
@@ -239,7 +239,7 @@ def get_parser():
"\n6 bwd+wrw"
)
parser.add_argument(
"-spatial_dim",
"--spatial_dim",
"-_",
default=2,
type=int,
@@ -247,7 +247,7 @@ def get_parser():
help="convolution spatial dimension (Default-2)"
)
parser.add_argument(
"-batchsize",
"--batchsize",
"-n",
default=100,
type=int,
@@ -255,7 +255,7 @@ def get_parser():
help="Mini-batch size (Default=100)"
)
parser.add_argument(
"-in_channels",
"--in_channels",
"-c",
default=3,
type=int,
@@ -263,7 +263,7 @@ def get_parser():
help="Number of Input Channels (Default=3)"
)
parser.add_argument(
"-in_d",
"--in_d",
"-!",
default=32,
type=int,
@@ -271,7 +271,7 @@ def get_parser():
help="Input Depth (Default=32)"
)
parser.add_argument(
"-in_h",
"--in_h",
"-H",
default=32,
type=int,
@@ -279,7 +279,7 @@ def get_parser():
help="Input Height (Default=32)"
)
parser.add_argument(
"-in_w",
"--in_w",
"-W",
default=32,
type=int,
@@ -287,7 +287,7 @@ def get_parser():
help="Input Width (Default=32)"
)
parser.add_argument(
"-out_channels",
"--out_channels",
"-k",
default=32,
type=int,
@@ -295,7 +295,7 @@ def get_parser():
help="Number of Output Channels (Default=32)"
)
parser.add_argument(
"-fil_d",
"--fil_d",
"-@",
default=3,
type=int,
@@ -303,7 +303,7 @@ def get_parser():
help="Filter Depth (Default=3)"
)
parser.add_argument(
"-fil_h",
"--fil_h",
"-y",
default=3,
type=int,
@@ -311,7 +311,7 @@ def get_parser():
help="Filter Height (Default=3)"
)
parser.add_argument(
"-fil_w",
"--fil_w",
"-x",
default=3,
type=int,
@@ -319,7 +319,7 @@ def get_parser():
help="Filter Width (Default=3)"
)
parser.add_argument(
"-conv_stride_d",
"--conv_stride_d",
"-#",
default=1,
type=int,
@@ -327,7 +327,7 @@ def get_parser():
help="Convolution Stride for Depth (Default=1)"
)
parser.add_argument(
"-conv_stride_h",
"--conv_stride_h",
"-u",
default=1,
type=int,
@@ -335,7 +335,7 @@ def get_parser():
help="Convolution Stride for Height (Default=1)"
)
parser.add_argument(
"-conv_stride_w",
"--conv_stride_w",
"-v",
default=1,
type=int,
@@ -343,7 +343,7 @@ def get_parser():
help="Convolution Stride for Width (Default=1)"
)
parser.add_argument(
"-pad_d",
"--pad_d",
"-$",
default=1,
type=int,
@@ -351,7 +351,7 @@ def get_parser():
help="Zero Padding for Depth (Default=0)"
)
parser.add_argument(
"-pad_h",
"--pad_h",
"-p",
default=1,
type=int,
@@ -359,7 +359,7 @@ def get_parser():
help="Zero Padding for Height (Default=0)"
)
parser.add_argument(
"-pad_w",
"--pad_w",
"-q",
default=1,
type=int,
@@ -367,7 +367,7 @@ def get_parser():
help="Zero Padding for Width (Default=0)"
)
parser.add_argument(
"-verify",
"--verify",
"-V",
default=1,
type=int,
@@ -375,7 +375,7 @@ def get_parser():
help="Verify Each Layer (Default=1)"
)
parser.add_argument(
"-time",
"--time",
"-t",
default=0,
type=int,
@@ -383,7 +383,7 @@ def get_parser():
help="Time Each Layer (Default=0)"
)
parser.add_argument(
"-dilation_d",
"--dilation_d",
"-^",
default=1,
type=int,
@@ -391,7 +391,7 @@ def get_parser():
help="Dilation of Filter Depth (Default=1)"
)
parser.add_argument(
"-dilation_h",
"--dilation_h",
"-l",
default=1,
type=int,
@@ -399,7 +399,7 @@ def get_parser():
help="Dilation of Filter Height (Default=1)"
)
parser.add_argument(
"-dilation_w",
"--dilation_w",
"-j",
default=1,
type=int,
@@ -407,7 +407,7 @@ def get_parser():
help="Dilation of Filter Width (Default=1)"
)
parser.add_argument(
"-group_count",
"--group_count",
"-g",
type=int,
default=1,

View File

@@ -0,0 +1,185 @@
#!/usr/bin/env python3
import sys
import os
import argparse
import pandas as pd
import csv
from convert_miopen_driver_to_profiler import get_parser, init_const_args, process_miopen_driver_name, \
get_ck_grouped_conv_fwd_cmd, get_ck_grouped_conv_bwd_data_cmd, get_ck_grouped_conv_bwd_weight_cmd
def parse_cli_args():
"""Parse command line arguments"""
parser = argparse.ArgumentParser(description="Run CK convolution profiler.")
parser.add_argument("--fremont-csv-file", type=str, dest="fremont_csv_file", required=True, help="Path to the CSV file containing Fremont test cases.")
parser.add_argument("--ktn-csv-file", type=str, dest="ktn_csv_file", required=True, help="Path to the CSV file containing KTN test cases.")
parser.add_argument("--fwd-only", action="store_true", help="Run only forward convolution.")
parser.add_argument("--bwd-data-only", action="store_true", help="Run only backward data convolution.")
parser.add_argument("--bwd-weight-only", action="store_true", help="Run only backward weight convolution.")
parser.add_argument("--no-verification", action="store_true", help="Disable verification in the CK profiler.")
parser.add_argument("--full-set", action="store_true", help="Create a full set of tests. By default only a subset of the raw data is used.")
parser.add_argument("--output-path", type=str, dest="output_path", default=".", help="Path to save the output files. Default is current directory.")
args, unknown_args = parser.parse_known_args()
if unknown_args:
print(f"Unknown arguments: {unknown_args}", file=sys.stderr)
sys.exit(1)
return args
def parse_profiler_command(args, fwd_only=False, bwd_data_only=False, bwd_weight_only=False):
# MIOpen get number of channel per all groups, CK profiler get number of
# channel per group
args.in_channels = int(args.in_channels / args.group_count)
args.out_channels = int(args.out_channels / args.group_count)
cmd = None
if fwd_only:
args.forw = 1
cmd = get_ck_grouped_conv_fwd_cmd(args)
if bwd_data_only:
args.forw = 2
cmd = get_ck_grouped_conv_bwd_data_cmd(args)
if bwd_weight_only:
args.forw = 4
cmd = get_ck_grouped_conv_bwd_weight_cmd(args)
return cmd
def parse_fremont_profiler_commands(csv_file, no_verification=False, fwd_only=False, bwd_data_only=False, bwd_weight_only=False):
if not os.path.isfile(csv_file):
print(f"Error: The specified CSV file '{csv_file}' does not exist.", file=sys.stderr)
sys.exit(1)
df = pd.read_csv(csv_file)
shapes = df['Shape'].tolist()
parser = get_parser()
commands = []
for i, line in enumerate(shapes):
try:
args, unknown = parser.parse_known_args(line.split())
init_const_args(args)
process_miopen_driver_name(args, unknown)
assert len(unknown) == 4 and unknown[0] == "--fil_layout" and unknown[2] == "--out_layout" and unknown[1] == unknown[3], \
f"Error: Unknown arguments do not match: {unknown}"
assert unknown[1] == args.in_layout, \
f"Error: Input layout does not match unknown arguments: {unknown[1]} != {args.in_layout}"
if no_verification:
args.verify = 0
# Ensure we run always the timing.
args.time = 1
command = parse_profiler_command(args,
fwd_only=fwd_only,
bwd_data_only=bwd_data_only,
bwd_weight_only=bwd_weight_only)
if command is not None:
commands.append(command)
except AttributeError as e:
print(f"Error processing line {i}: {line}. Skipping the line.")
continue
return commands
def process_miopen_driver(args, unknown):
if "convint8" in unknown:
args.data_type = 'int8'
elif "convbfp16" in unknown:
args.data_type = 'bfp16'
elif "convfp16" in unknown:
args.data_type = 'fp16'
elif "conv" in unknown:
args.data_type = 'fp32'
else:
print('Not supported driver (supported: conv, convfp16, convint8,'
' convbfp16).')
exit(1)
def parse_ktn_command(csv_file, no_verification=False, fwd_only=False, bwd_data_only=False, bwd_weight_only=False):
if not os.path.isfile(csv_file):
print(f"Error: The specified CSV file '{csv_file}' does not exist.", file=sys.stderr)
sys.exit(1)
df = pd.read_csv(csv_file)
# Remove the KTN commands where column 'Group Size' has value 1
df = df[df['Group Size'] != 1]
commands = []
parser = get_parser()
for i, cmd in enumerate(df['Command']):
cmd = cmd.strip()
args, _ = parser.parse_known_args(cmd)
init_const_args(args)
process_miopen_driver(args, cmd.split()[0])
if no_verification:
args.verify = 0
else:
args.verify = 1
# Ensure we run always the timing.
args.time = 1
command = parse_profiler_command(args,
fwd_only=fwd_only,
bwd_data_only=bwd_data_only,
bwd_weight_only=bwd_weight_only)
if command is not None:
commands.append(command)
return commands
def main():
args = parse_cli_args()
# Initialize random seed for reproducibility
seed = 42
n_fremont_shapes = 1000
n_ktn_shapes = 1000
fremont_commands = parse_fremont_profiler_commands(args.fremont_csv_file,
no_verification=args.no_verification,
fwd_only=args.fwd_only,
bwd_data_only=args.bwd_data_only,
bwd_weight_only=args.bwd_weight_only)
ktn_commands = parse_ktn_command(args.ktn_csv_file,
no_verification=args.no_verification,
fwd_only=args.fwd_only,
bwd_data_only=args.bwd_data_only,
bwd_weight_only=args.bwd_weight_only)
# Create a DataFrame to hold the commands
commands_fremont_df = pd.DataFrame({
'Command': fremont_commands,
})
commands_ktn_df = pd.DataFrame({
'Command': ktn_commands,
})
# Take a randomly sampled subset of Fremont commands
if not args.full_set:
commands_fremont_df = commands_fremont_df.sample(n=min(n_fremont_shapes, len(commands_fremont_df)), random_state=seed)
commands_ktn_df = commands_ktn_df.sample(n=min(n_ktn_shapes, len(commands_ktn_df)), random_state=seed)
output_file = os.path.join(args.output_path, "ck_profiler_commands.csv")
with open(output_file, "w") as f:
csv_writer = csv.writer(f, delimiter=',', quotechar='"', quoting=csv.QUOTE_MINIMAL)
for command in commands_fremont_df['Command']:
csv_writer.writerow(command)
for command in commands_ktn_df['Command']:
csv_writer.writerow(command)
print(f"Commands saved to {output_file}")
if __name__ == "__main__":
main()

View File

@@ -4,20 +4,11 @@ import os
import argparse
import subprocess
import sys
import pandas as pd
from convert_miopen_driver_to_profiler import get_parser, init_const_args, process_miopen_driver_name, \
get_ck_grouped_conv_fwd_cmd, get_ck_grouped_conv_bwd_data_cmd, get_ck_grouped_conv_bwd_weight_cmd
def parse_cli_args():
"""Parse command line arguments"""
parser = argparse.ArgumentParser(description="Run CK convolution profiler.")
parser.add_argument("--csv-file", type=str, dest="csv_file", required=True, help="Path to the CSV file containing test cases.")
parser.add_argument("--fwd-only", action="store_true", help="Run only forward convolution.")
parser.add_argument("--bwd-data-only", action="store_true", help="Run only backward data convolution.")
parser.add_argument("--bwd-weight-only", action="store_true", help="Run only backward weight convolution.")
parser.add_argument("--start", type=int, default=None, help="Start index for processing shapes in the CSV file.")
parser.add_argument("--end", type=int, default=None, help="End index for processing shapes in the CSV file. If None, process all shapes.")
parser.add_argument("--no-verification", action="store_true", help="Disable verification in the CK profiler.")
parser.add_argument("--log-to-stdout", action="store_true", help="Log profiler output to stdout instead of /dev/null.")
args, unknown_args = parser.parse_known_args()
@@ -28,65 +19,6 @@ def parse_cli_args():
return args
def parse_profiler_command(args, fwd_only=False, bwd_data_only=False, bwd_weight_only=False):
# MIOpen get number of channel per all groups, CK profiler get number of
# channel per group
args.in_channels = int(args.in_channels / args.group_count)
args.out_channels = int(args.out_channels / args.group_count)
cmd = None
if fwd_only:
args.forw = 1
cmd = get_ck_grouped_conv_fwd_cmd(args)
if bwd_data_only:
args.forw = 2
cmd = get_ck_grouped_conv_bwd_data_cmd(args)
if bwd_weight_only:
args.forw = 4
cmd = get_ck_grouped_conv_bwd_weight_cmd(args)
return cmd
def get_profiler_commands(csv_file, no_verification=False, fwd_only=False, bwd_data_only=False, bwd_weight_only=False):
if not os.path.isfile(csv_file):
print(f"Error: The specified CSV file '{csv_file}' does not exist.", file=sys.stderr)
sys.exit(1)
df = pd.read_csv(csv_file)
shapes = df['Shape'].tolist()
parser = get_parser()
commands = []
for i, line in enumerate(shapes):
try:
args, unknown = parser.parse_known_args(line.split())
init_const_args(args)
process_miopen_driver_name(args, unknown)
assert len(unknown) == 4 and unknown[0] == "--fil_layout" and unknown[2] == "--out_layout" and unknown[1] == unknown[3], \
f"Error: Unknown arguments do not match: {unknown}"
assert unknown[1] == args.in_layout, \
f"Error: Input layout does not match unknown arguments: {unknown[1]} != {args.in_layout}"
if no_verification:
args.verify = 0
# Ensure we run always the timing.
args.time = 1
command = parse_profiler_command(args,
fwd_only=fwd_only,
bwd_data_only=bwd_data_only,
bwd_weight_only=bwd_weight_only)
if command is not None:
commands.append(command)
except AttributeError as e:
print(f"Error processing line {i}: {line}. Skipping the line.")
continue
return commands
def run_ck_profiler_cmd(cmd, log_to_stdout=False):
cmd_concatenated_str = ""
for arg in cmd: