From 3b8dfae5bc3c571cd6fd813d58fde8a5dd7087ad Mon Sep 17 00:00:00 2001 From: =?UTF-8?q?Ville=20Pietil=C3=A4?= <> Date: Fri, 13 Jun 2025 13:49:38 +0000 Subject: [PATCH] Add script to generate CK profiler commands from the KTN data. --- script/convert_miopen_driver_to_profiler.py | 46 ++--- script/generate_test_data.py | 185 ++++++++++++++++++++ script/run_conv_profiler.py | 68 ------- 3 files changed, 208 insertions(+), 91 deletions(-) create mode 100644 script/generate_test_data.py diff --git a/script/convert_miopen_driver_to_profiler.py b/script/convert_miopen_driver_to_profiler.py index 940dcce201..d8b4422bca 100644 --- a/script/convert_miopen_driver_to_profiler.py +++ b/script/convert_miopen_driver_to_profiler.py @@ -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, diff --git a/script/generate_test_data.py b/script/generate_test_data.py new file mode 100644 index 0000000000..d018313383 --- /dev/null +++ b/script/generate_test_data.py @@ -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() \ No newline at end of file diff --git a/script/run_conv_profiler.py b/script/run_conv_profiler.py index 3a9f488ddf..6e78add8a9 100644 --- a/script/run_conv_profiler.py +++ b/script/run_conv_profiler.py @@ -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: