ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-05-14 02:02:46 +00:00
Code Issues Packages Projects Releases Wiki Activity

Update pre-commit to fixed versions, run remod for ck_tile (#2895)

Browse Source 
* Fix ruff linter errors

* Fix remod dos2unix command

* Clang format

* Ignore utility in remod

* Run remod

* Specify clang-format version in pre-commit

* Specify ruff version

* Include PoolKernelArgs in reference_pool

* Add calculate_total_elements to reference batched contraction

* Fix calculate_total_elements declaration

* Refactor remod pre-commit hook

* Fix Aquant tests

---------

Co-authored-by: Illia Silin <98187287+illsilin@users.noreply.github.com>

[ROCm/composable_kernel commit: d40b50b9d5]
This commit is contained in:
Johannes Graner
2025-10-17 00:29:17 +02:00
committed by GitHub
parent 6066662785
commit 580a54b400
77 changed files with 21671 additions and 9858 deletions
Download Patch File Download Diff File Expand all files Collapse all files

212
test_data/generate_model_configs.py
View File

@@ -10,28 +10,37 @@ and saves them as CSV files that can be read by the shell script.
"""
import csv
import itertools
import argparse
def generate_2d_configs(mode='full'):
def generate_2d_configs(mode="full"):
"""Generate all 2D model configuration combinations
Args:
mode: 'small' for minimal set (~50 configs), 'half' for reduced set (~250 configs), 'full' for comprehensive set (~500 configs)
"""
# Define parameter ranges
models_2d = [
'resnet18', 'resnet34', 'resnet50',
'mobilenet_v2', 'mobilenet_v3_large', 'mobilenet_v3_small',
'vgg11', 'vgg16', 'vgg19',
'alexnet', 'googlenet',
'densenet121', 'densenet161',
'squeezenet1_0', 'squeezenet1_1',
'shufflenet_v2_x1_0'
"resnet18",
"resnet34",
"resnet50",
"mobilenet_v2",
"mobilenet_v3_large",
"mobilenet_v3_small",
"vgg11",
"vgg16",
"vgg19",
"alexnet",
"googlenet",
"densenet121",
"densenet161",
"squeezenet1_0",
"squeezenet1_1",
"shufflenet_v2_x1_0",
]
if mode == 'small':
if mode == "small":
# Minimal set for quick testing
batch_sizes = [1, 8] # Just two batch sizes
# Very limited input dimensions - only 2 key sizes
@@ -41,12 +50,12 @@ def generate_2d_configs(mode='full'):
]
# Use only first 3 models for minimal testing
models_2d = models_2d[:3] # Only resnet18, resnet34, resnet50
elif mode == 'half':
elif mode == "half":
# Reduced set for faster testing
batch_sizes = [1, 8, 32] # Small, medium, large
# Reduced input dimensions - 5 key sizes
input_dims = [
(64, 64), # Small
(64, 64), # Small
(224, 224), # Standard (most common)
(512, 512), # Large
(224, 320), # Rectangular
@@ -57,18 +66,23 @@ def generate_2d_configs(mode='full'):
batch_sizes = [1, 4, 8, 16, 32]
# More dimensions but skip some redundant ones
input_dims = [
(64, 64), (128, 128), (224, 224), (256, 256), (512, 512), # Square
(224, 320), (320, 224), # Rectangular (reduced from 4)
(64, 64),
(128, 128),
(224, 224),
(256, 256),
(512, 512), # Square
(224, 320),
(320, 224), # Rectangular (reduced from 4)
(227, 227), # AlexNet preferred
(299, 299) # Inception preferred
(299, 299), # Inception preferred
]
precisions = ['fp32'] #, 'fp16', 'bf16']
precisions = ["fp32"] # , 'fp16', 'bf16']
channels = [3] # Most models expect RGB
configs = []
config_id = 1
# Generate all combinations (but limit to reasonable subset)
for model in models_2d:
for batch_size in batch_sizes:
@@ -77,36 +91,37 @@ def generate_2d_configs(mode='full'):
# Skip some combinations to keep dataset manageable
if batch_size > 16 and height > 256:
continue # Skip large batch + large image combinations
if precision != 'fp32' and batch_size < 8:
if precision != "fp32" and batch_size < 8:
continue # Skip mixed precision with tiny batches
config_name = f"{model}_b{batch_size}_{height}x{width}_{precision}"
config = {
'config_name': config_name,
'model': model,
'batch_size': batch_size,
'channels': channels[0],
'height': height,
'width': width,
'precision': precision
"config_name": config_name,
"model": model,
"batch_size": batch_size,
"channels": channels[0],
"height": height,
"width": width,
"precision": precision,
}
configs.append(config)
config_id += 1
return configs
def generate_3d_configs(mode='full'):
def generate_3d_configs(mode="full"):
"""Generate all 3D model configuration combinations
Args:
mode: 'small' for minimal set (~10 configs), 'half' for reduced set (~50 configs), 'full' for comprehensive set (~100 configs)
"""
models_3d = ['r3d_18', 'mc3_18', 'r2plus1d_18']
if mode == 'small':
models_3d = ["r3d_18", "mc3_18", "r2plus1d_18"]
if mode == "small":
# Minimal set for quick testing
batch_sizes = [1, 4] # Just two batch sizes
temporal_sizes = [8] # Only smallest temporal size
@@ -116,7 +131,7 @@ def generate_3d_configs(mode='full'):
]
# Use only first model for minimal testing
models_3d = models_3d[:1] # Only r3d_18
elif mode == 'half':
elif mode == "half":
# Reduced set for faster testing
batch_sizes = [1, 4, 8] # Skip batch_size=2
temporal_sizes = [8, 16] # Skip 32 (most expensive)
@@ -124,7 +139,7 @@ def generate_3d_configs(mode='full'):
input_dims = [
(112, 112), # Small (common for video)
(224, 224), # Standard
(224, 320) # Rectangular
(224, 320), # Rectangular
]
else: # full mode
# More comprehensive but still reasonable
@@ -132,15 +147,18 @@ def generate_3d_configs(mode='full'):
temporal_sizes = [8, 16, 32]
# More dimensions
input_dims = [
(112, 112), (224, 224), (256, 256), # Standard sizes
(224, 320), (320, 224) # Rectangular
(112, 112),
(224, 224),
(256, 256), # Standard sizes
(224, 320),
(320, 224), # Rectangular
]
precisions = ['fp32'] #, 'fp16'] # Skip bf16 for 3D to reduce combinations
precisions = ["fp32"] # , 'fp16'] # Skip bf16 for 3D to reduce combinations
channels = [3]
configs = []
for model in models_3d:
for batch_size in batch_sizes:
for temporal_size in temporal_sizes:
@@ -151,75 +169,97 @@ def generate_3d_configs(mode='full'):
continue
if batch_size > 2 and height > 224:
continue
config_name = f"{model}_b{batch_size}_t{temporal_size}_{height}x{width}_{precision}"
config = {
'config_name': config_name,
'model': model,
'batch_size': batch_size,
'channels': channels[0],
'temporal_size': temporal_size,
'height': height,
'width': width,
'precision': precision
}
"config_name": config_name,
"model": model,
"batch_size": batch_size,
"channels": channels[0],
"temporal_size": temporal_size,
"height": height,
"width": width,
"precision": precision,
}
configs.append(config)
return configs
def save_configs_to_csv(configs, filename, config_type):
"""Save configurations to CSV file"""
if not configs:
print(f"No {config_type} configurations generated")
return
fieldnames = list(configs[0].keys())
with open(filename, 'w', newline='\n', encoding='utf-8') as csvfile:
with open(filename, "w", newline="\n", encoding="utf-8") as csvfile:
csvfile.write(f"# {config_type} Model Configurations\n")
csvfile.write(f"# Generated {len(configs)} configurations\n")
writer = csv.DictWriter(csvfile, fieldnames=fieldnames, lineterminator='\n')
writer = csv.DictWriter(csvfile, fieldnames=fieldnames, lineterminator="\n")
writer.writeheader()
for config in configs:
writer.writerow(config)
print(f"Generated {len(configs)} {config_type} configurations → {filename}")
def main():
parser = argparse.ArgumentParser(description='Generate model configuration combinations')
parser.add_argument('--output-2d', type=str, default='model_configs_2d.csv',
help='Output file for 2D configurations')
parser.add_argument('--output-3d', type=str, default='model_configs_3d.csv',
help='Output file for 3D configurations')
parser.add_argument('--mode', choices=['small', 'half', 'full'], default='full',
help='Configuration mode: small (~60 total), half (~300 total) or full (~600 total) (default: half)')
parser.add_argument('--limit', type=int,
help='Limit number of configurations per type (for testing)')
parser = argparse.ArgumentParser(
description="Generate model configuration combinations"
)
parser.add_argument(
"--output-2d",
type=str,
default="model_configs_2d.csv",
help="Output file for 2D configurations",
)
parser.add_argument(
"--output-3d",
type=str,
default="model_configs_3d.csv",
help="Output file for 3D configurations",
)
parser.add_argument(
"--mode",
choices=["small", "half", "full"],
default="full",
help="Configuration mode: small (~60 total), half (~300 total) or full (~600 total) (default: half)",
)
parser.add_argument(
"--limit",
type=int,
help="Limit number of configurations per type (for testing)",
)
args = parser.parse_args()
print(f"Generating {args.mode} model configurations...")
print("Generating 2D model configurations...")
configs_2d = generate_2d_configs(mode=args.mode)
if args.limit:
configs_2d = configs_2d[:args.limit]
configs_2d = configs_2d[: args.limit]
save_configs_to_csv(configs_2d, args.output_2d, "2D")
print("Generating 3D model configurations...")
configs_3d = generate_3d_configs(mode=args.mode)
if args.limit:
configs_3d = configs_3d[:args.limit]
configs_3d = configs_3d[: args.limit]
save_configs_to_csv(configs_3d, args.output_3d, "3D")
print(f"\nTotal configurations: {len(configs_2d)} 2D + {len(configs_3d)} 3D = {len(configs_2d) + len(configs_3d)}")
print(
f"\nTotal configurations: {len(configs_2d)} 2D + {len(configs_3d)} 3D = {len(configs_2d) + len(configs_3d)}"
)
print("\nTo use these configurations:")
print(" Update generate_test_dataset.sh to read from these CSV files")
if __name__ == "__main__":
main()

530
test_data/miopen_to_csv.py
View File

@@ -18,301 +18,428 @@ import csv
import re
import os
def parse_miopen_command(command_line):
"""
Parse MIOpen driver command line into parameter dictionary
Example input:
./bin/MIOpenDriver conv -n 4 -c 3 -H 224 -W 224 -k 64 -y 3 -x 3 -p 1 -q 1 -u 1 -v 1 -l 1 -j 1 -m conv -g 1 -F 1 -t 1
Returns dict with parsed parameters or None if parsing fails
"""
if not command_line.strip().startswith('./bin/MIOpenDriver conv'):
if not command_line.strip().startswith("./bin/MIOpenDriver conv"):
return None
# Extract parameters using regex
params = {}
# Parameter mapping: flag -> description
# Support both short (-D) and long (--in_d) parameter formats
param_patterns = {
'n': r'-n\s+(\d+)', # batch size
'c': r'-c\s+(\d+)', # input channels
'k': r'-k\s+(\d+)', # output channels
'H': r'-H\s+(\d+)', # input height
'W': r'-W\s+(\d+)', # input width
'D': r'(?:-D|--in_d)\s+(\d+)', # input depth (3D only) - supports both -D and --in_d
'y': r'-y\s+(\d+)', # kernel height
'x': r'-x\s+(\d+)', # kernel width
'z': r'(?:-z|--fil_d)\s+(\d+)', # kernel depth (3D only) - supports both -z and --fil_d
'u': r'-u\s+(\d+)', # stride height
'v': r'-v\s+(\d+)', # stride width
'w': r'(?:-w|--conv_stride_d)\s+(\d+)', # stride depth (3D only) - supports both -w and --conv_stride_d
'p': r'-p\s+(\d+)', # pad height
'q': r'-q\s+(\d+)', # pad width
's': r'(?:-s|--pad_d)\s+(\d+)', # pad depth (3D only) - supports both -s and --pad_d
'l': r'-l\s+(\d+)', # dilation height
'j': r'-j\s+(\d+)', # dilation width
'r': r'(?:-r|--dilation_d)\s+(\d+)', # dilation depth (3D only) - supports both -r and --dilation_d
'g': r'-g\s+(\d+)', # groups
'F': r'-F\s+(\d+)', # direction (1=fwd, 2=bwd_weight, 4=bwd_data)
"n": r"-n\s+(\d+)", # batch size
"c": r"-c\s+(\d+)", # input channels
"k": r"-k\s+(\d+)", # output channels
"H": r"-H\s+(\d+)", # input height
"W": r"-W\s+(\d+)", # input width
"D": r"(?:-D|--in_d)\s+(\d+)", # input depth (3D only) - supports both -D and --in_d
"y": r"-y\s+(\d+)", # kernel height
"x": r"-x\s+(\d+)", # kernel width
"z": r"(?:-z|--fil_d)\s+(\d+)", # kernel depth (3D only) - supports both -z and --fil_d
"u": r"-u\s+(\d+)", # stride height
"v": r"-v\s+(\d+)", # stride width
"w": r"(?:-w|--conv_stride_d)\s+(\d+)", # stride depth (3D only) - supports both -w and --conv_stride_d
"p": r"-p\s+(\d+)", # pad height
"q": r"-q\s+(\d+)", # pad width
"s": r"(?:-s|--pad_d)\s+(\d+)", # pad depth (3D only) - supports both -s and --pad_d
"l": r"-l\s+(\d+)", # dilation height
"j": r"-j\s+(\d+)", # dilation width
"r": r"(?:-r|--dilation_d)\s+(\d+)", # dilation depth (3D only) - supports both -r and --dilation_d
"g": r"-g\s+(\d+)", # groups
"F": r"-F\s+(\d+)", # direction (1=fwd, 2=bwd_weight, 4=bwd_data)
}
for param, pattern in param_patterns.items():
match = re.search(pattern, command_line)
if match:
params[param] = int(match.group(1))
return params if params else None
def miopen_to_conv_param(miopen_params):
"""
Convert MIOpen parameters to CK ConvParam format
Returns dictionary in CSV format or None if conversion fails
"""
if not miopen_params:
return None
# Determine if 2D or 3D convolution
is_3d = 'D' in miopen_params or 'z' in miopen_params or 'w' in miopen_params or 'r' in miopen_params or 's' in miopen_params
is_3d = (
"D" in miopen_params
or "z" in miopen_params
or "w" in miopen_params
or "r" in miopen_params
or "s" in miopen_params
)
# Extract basic parameters with defaults
ndim = 3 if is_3d else 2
groups = miopen_params.get('g', 1)
batch_size = miopen_params.get('n', 1)
groups = miopen_params.get("g", 1)
batch_size = miopen_params.get("n", 1)
# MIOpen uses total channels (C*G), CK uses channels per group
out_channels_total = miopen_params.get('k', 64)
in_channels_total = miopen_params.get('c', 3)
out_channels_total = miopen_params.get("k", 64)
in_channels_total = miopen_params.get("c", 3)
out_channels = out_channels_total // groups # CK format: channels per group
in_channels = in_channels_total // groups # CK format: channels per group
in_channels = in_channels_total // groups # CK format: channels per group
if is_3d:
# 3D convolution
kernel_d = miopen_params.get('z', 3)
kernel_h = miopen_params.get('y', 3)
kernel_w = miopen_params.get('x', 3)
input_d = miopen_params.get('D', 16)
input_h = miopen_params.get('H', 32)
input_w = miopen_params.get('W', 32)
stride_d = miopen_params.get('w', 1)
stride_h = miopen_params.get('u', 1)
stride_w = miopen_params.get('v', 1)
dilation_d = miopen_params.get('r', 1)
dilation_h = miopen_params.get('l', 1)
dilation_w = miopen_params.get('j', 1)
pad_d = miopen_params.get('s', 0)
pad_h = miopen_params.get('p', 0)
pad_w = miopen_params.get('q', 0)
kernel_d = miopen_params.get("z", 3)
kernel_h = miopen_params.get("y", 3)
kernel_w = miopen_params.get("x", 3)
input_d = miopen_params.get("D", 16)
input_h = miopen_params.get("H", 32)
input_w = miopen_params.get("W", 32)
stride_d = miopen_params.get("w", 1)
stride_h = miopen_params.get("u", 1)
stride_w = miopen_params.get("v", 1)
dilation_d = miopen_params.get("r", 1)
dilation_h = miopen_params.get("l", 1)
dilation_w = miopen_params.get("j", 1)
pad_d = miopen_params.get("s", 0)
pad_h = miopen_params.get("p", 0)
pad_w = miopen_params.get("q", 0)
# Calculate output dimensions
output_d = (input_d + 2 * pad_d - dilation_d * (kernel_d - 1) - 1) // stride_d + 1
output_h = (input_h + 2 * pad_h - dilation_h * (kernel_h - 1) - 1) // stride_h + 1
output_w = (input_w + 2 * pad_w - dilation_w * (kernel_w - 1) - 1) // stride_w + 1
output_d = (
input_d + 2 * pad_d - dilation_d * (kernel_d - 1) - 1
) // stride_d + 1
output_h = (
input_h + 2 * pad_h - dilation_h * (kernel_h - 1) - 1
) // stride_h + 1
output_w = (
input_w + 2 * pad_w - dilation_w * (kernel_w - 1) - 1
) // stride_w + 1
# Skip invalid configurations
if output_d <= 0 or output_h <= 0 or output_w <= 0:
return None
direction = miopen_params.get('F', 1) # 1=fwd, 2=bwd_weight, 4=bwd_data
direction_name = {1: 'fwd', 2: 'bwd_weight', 4: 'bwd_data'}.get(direction, 'fwd')
direction = miopen_params.get("F", 1) # 1=fwd, 2=bwd_weight, 4=bwd_data
direction_name = {1: "fwd", 2: "bwd_weight", 4: "bwd_data"}.get(
direction, "fwd"
)
return {
'NDim': ndim,
'Groups': groups,
'BatchSize': batch_size,
'OutChannels': out_channels,
'InChannels': in_channels,
'KernelD': kernel_d, 'KernelH': kernel_h, 'KernelW': kernel_w,
'InputD': input_d, 'InputH': input_h, 'InputW': input_w,
'OutputD': output_d, 'OutputH': output_h, 'OutputW': output_w,
'StrideD': stride_d, 'StrideH': stride_h, 'StrideW': stride_w,
'DilationD': dilation_d, 'DilationH': dilation_h, 'DilationW': dilation_w,
'LeftPadD': pad_d, 'LeftPadH': pad_h, 'LeftPadW': pad_w,
'RightPadD': pad_d, 'RightPadH': pad_h, 'RightPadW': pad_w,
'TestName': f'MIOpen_3D_{direction_name}'
"NDim": ndim,
"Groups": groups,
"BatchSize": batch_size,
"OutChannels": out_channels,
"InChannels": in_channels,
"KernelD": kernel_d,
"KernelH": kernel_h,
"KernelW": kernel_w,
"InputD": input_d,
"InputH": input_h,
"InputW": input_w,
"OutputD": output_d,
"OutputH": output_h,
"OutputW": output_w,
"StrideD": stride_d,
"StrideH": stride_h,
"StrideW": stride_w,
"DilationD": dilation_d,
"DilationH": dilation_h,
"DilationW": dilation_w,
"LeftPadD": pad_d,
"LeftPadH": pad_h,
"LeftPadW": pad_w,
"RightPadD": pad_d,
"RightPadH": pad_h,
"RightPadW": pad_w,
"TestName": f"MIOpen_3D_{direction_name}",
}
else:
# 2D convolution
kernel_h = miopen_params.get('y', 3)
kernel_w = miopen_params.get('x', 3)
input_h = miopen_params.get('H', 32)
input_w = miopen_params.get('W', 32)
stride_h = miopen_params.get('u', 1)
stride_w = miopen_params.get('v', 1)
dilation_h = miopen_params.get('l', 1)
dilation_w = miopen_params.get('j', 1)
pad_h = miopen_params.get('p', 0)
pad_w = miopen_params.get('q', 0)
kernel_h = miopen_params.get("y", 3)
kernel_w = miopen_params.get("x", 3)
input_h = miopen_params.get("H", 32)
input_w = miopen_params.get("W", 32)
stride_h = miopen_params.get("u", 1)
stride_w = miopen_params.get("v", 1)
dilation_h = miopen_params.get("l", 1)
dilation_w = miopen_params.get("j", 1)
pad_h = miopen_params.get("p", 0)
pad_w = miopen_params.get("q", 0)
# Calculate output dimensions
output_h = (input_h + 2 * pad_h - dilation_h * (kernel_h - 1) - 1) // stride_h + 1
output_w = (input_w + 2 * pad_w - dilation_w * (kernel_w - 1) - 1) // stride_w + 1
output_h = (
input_h + 2 * pad_h - dilation_h * (kernel_h - 1) - 1
) // stride_h + 1
output_w = (
input_w + 2 * pad_w - dilation_w * (kernel_w - 1) - 1
) // stride_w + 1
# Skip invalid configurations
if output_h <= 0 or output_w <= 0:
return None
direction = miopen_params.get('F', 1)
direction_name = {1: 'fwd', 2: 'bwd_weight', 4: 'bwd_data'}.get(direction, 'fwd')
direction = miopen_params.get("F", 1)
direction_name = {1: "fwd", 2: "bwd_weight", 4: "bwd_data"}.get(
direction, "fwd"
)
return {
'NDim': ndim,
'Groups': groups,
'BatchSize': batch_size,
'OutChannels': out_channels,
'InChannels': in_channels,
'KernelH': kernel_h, 'KernelW': kernel_w,
'InputH': input_h, 'InputW': input_w,
'OutputH': output_h, 'OutputW': output_w,
'StrideH': stride_h, 'StrideW': stride_w,
'DilationH': dilation_h, 'DilationW': dilation_w,
'LeftPadH': pad_h, 'LeftPadW': pad_w,
'RightPadH': pad_h, 'RightPadW': pad_w,
'TestName': f'MIOpen_2D_{direction_name}'
"NDim": ndim,
"Groups": groups,
"BatchSize": batch_size,
"OutChannels": out_channels,
"InChannels": in_channels,
"KernelH": kernel_h,
"KernelW": kernel_w,
"InputH": input_h,
"InputW": input_w,
"OutputH": output_h,
"OutputW": output_w,
"StrideH": stride_h,
"StrideW": stride_w,
"DilationH": dilation_h,
"DilationW": dilation_w,
"LeftPadH": pad_h,
"LeftPadW": pad_w,
"RightPadH": pad_h,
"RightPadW": pad_w,
"TestName": f"MIOpen_2D_{direction_name}",
}
def write_csv_cases(test_cases, output_file, ndim):
"""Write test cases to CSV file"""
if not test_cases:
print(f"No {ndim}D test cases to write")
return
print(f"Writing {len(test_cases)} {ndim}D test cases to {output_file}")
# Define CSV headers based on dimension
if ndim == 2:
headers = ['NDim', 'Groups', 'BatchSize', 'OutChannels', 'InChannels',
'KernelH', 'KernelW', 'InputH', 'InputW', 'OutputH', 'OutputW',
'StrideH', 'StrideW', 'DilationH', 'DilationW',
'LeftPadH', 'LeftPadW', 'RightPadH', 'RightPadW', 'TestName']
headers = [
"NDim",
"Groups",
"BatchSize",
"OutChannels",
"InChannels",
"KernelH",
"KernelW",
"InputH",
"InputW",
"OutputH",
"OutputW",
"StrideH",
"StrideW",
"DilationH",
"DilationW",
"LeftPadH",
"LeftPadW",
"RightPadH",
"RightPadW",
"TestName",
]
else: # 3D
headers = ['NDim', 'Groups', 'BatchSize', 'OutChannels', 'InChannels',
'KernelD', 'KernelH', 'KernelW', 'InputD', 'InputH', 'InputW',
'OutputD', 'OutputH', 'OutputW', 'StrideD', 'StrideH', 'StrideW',
'DilationD', 'DilationH', 'DilationW',
'LeftPadD', 'LeftPadH', 'LeftPadW', 'RightPadD', 'RightPadH', 'RightPadW', 'TestName']
with open(output_file, 'w', newline='') as csvfile:
headers = [
"NDim",
"Groups",
"BatchSize",
"OutChannels",
"InChannels",
"KernelD",
"KernelH",
"KernelW",
"InputD",
"InputH",
"InputW",
"OutputD",
"OutputH",
"OutputW",
"StrideD",
"StrideH",
"StrideW",
"DilationD",
"DilationH",
"DilationW",
"LeftPadD",
"LeftPadH",
"LeftPadW",
"RightPadD",
"RightPadH",
"RightPadW",
"TestName",
]
with open(output_file, "w", newline="") as csvfile:
# Write header comment
csvfile.write(f"# {ndim}D Convolution Test Cases from MIOpen Commands\n")
csvfile.write(f"# Generated {len(test_cases)} test cases\n")
writer = csv.DictWriter(csvfile, fieldnames=headers)
writer.writeheader()
for test_case in test_cases:
# Only write fields that exist in headers
filtered_case = {k: v for k, v in test_case.items() if k in headers}
writer.writerow(filtered_case)
def main():
parser = argparse.ArgumentParser(description='Convert MIOpen commands to CSV test cases')
parser.add_argument('--input', type=str, required=True,
help='Input file with MIOpen driver commands')
parser.add_argument('--output', type=str,
help='Output CSV file (for mixed 2D/3D cases)')
parser.add_argument('--output-2d', type=str, default='miopen_conv_2d.csv',
help='Output CSV file for 2D cases')
parser.add_argument('--output-3d', type=str, default='miopen_conv_3d.csv',
help='Output CSV file for 3D cases')
parser.add_argument('--filter-duplicates', action='store_true',
help='Remove duplicate test cases')
parser.add_argument('--model-name', type=str, default='MIOpen',
help='Model name to use in test case names (default: MIOpen)')
parser = argparse.ArgumentParser(
description="Convert MIOpen commands to CSV test cases"
)
parser.add_argument(
"--input",
type=str,
required=True,
help="Input file with MIOpen driver commands",
)
parser.add_argument(
"--output", type=str, help="Output CSV file (for mixed 2D/3D cases)"
)
parser.add_argument(
"--output-2d",
type=str,
default="miopen_conv_2d.csv",
help="Output CSV file for 2D cases",
)
parser.add_argument(
"--output-3d",
type=str,
default="miopen_conv_3d.csv",
help="Output CSV file for 3D cases",
)
parser.add_argument(
"--filter-duplicates", action="store_true", help="Remove duplicate test cases"
)
parser.add_argument(
"--model-name",
type=str,
default="MIOpen",
help="Model name to use in test case names (default: MIOpen)",
)
args = parser.parse_args()
if not os.path.exists(args.input):
print(f"ERROR: Input file not found: {args.input}")
return 1
print(f"Parsing MIOpen commands from {args.input}...")
test_cases_2d = []
test_cases_3d = []
total_lines = 0
parsed_lines = 0
with open(args.input, 'r') as f:
with open(args.input, "r") as f:
for line_num, line in enumerate(f, 1):
total_lines += 1
line = line.strip()
# Skip empty lines and non-MIOpen commands
# Handle both direct commands and logged commands with MIOpen prefix
if not line:
continue
# Extract the actual MIOpenDriver command from logged format
if 'MIOpenDriver conv' in line:
if "MIOpenDriver conv" in line:
# Extract command after finding MIOpenDriver
command_start = line.find('./bin/MIOpenDriver conv')
command_start = line.find("./bin/MIOpenDriver conv")
if command_start != -1:
line = line[command_start:]
else:
# Handle cases where path might be different - create standard format
driver_start = line.find('MIOpenDriver conv')
driver_start = line.find("MIOpenDriver conv")
if driver_start != -1:
line = './bin/' + line[driver_start:]
line = "./bin/" + line[driver_start:]
else:
continue
elif not line.startswith('./bin/MIOpenDriver conv'):
elif not line.startswith("./bin/MIOpenDriver conv"):
continue
try:
# Parse MIOpen command
miopen_params = parse_miopen_command(line)
if not miopen_params:
continue
# Convert to ConvParam format
conv_param = miopen_to_conv_param(miopen_params)
if not conv_param:
continue
# Add model name to test name
conv_param['TestName'] = f"{args.model_name}_{conv_param['NDim']}D_fwd"
conv_param["TestName"] = f"{args.model_name}_{conv_param['NDim']}D_fwd"
# Separate 2D and 3D cases
if conv_param['NDim'] == 2:
if conv_param["NDim"] == 2:
test_cases_2d.append(conv_param)
else:
test_cases_3d.append(conv_param)
parsed_lines += 1
except Exception as e:
print(f"WARNING: Failed to parse line {line_num}: {e}")
continue
print(f"Processed {total_lines} lines, parsed {parsed_lines} commands")
print(f"Found {len(test_cases_2d)} 2D cases, {len(test_cases_3d)} 3D cases")
# Remove duplicates if requested
if args.filter_duplicates:
# Simple duplicate removal based on key parameters
def make_key(case):
if case['NDim'] == 2:
return (case['Groups'], case['BatchSize'], case['OutChannels'], case['InChannels'],
case['KernelH'], case['KernelW'], case['InputH'], case['InputW'],
case['StrideH'], case['StrideW'])
if case["NDim"] == 2:
return (
case["Groups"],
case["BatchSize"],
case["OutChannels"],
case["InChannels"],
case["KernelH"],
case["KernelW"],
case["InputH"],
case["InputW"],
case["StrideH"],
case["StrideW"],
)
else:
return (case['Groups'], case['BatchSize'], case['OutChannels'], case['InChannels'],
case['KernelD'], case['KernelH'], case['KernelW'],
case['InputD'], case['InputH'], case['InputW'],
case['StrideD'], case['StrideH'], case['StrideW'])
return (
case["Groups"],
case["BatchSize"],
case["OutChannels"],
case["InChannels"],
case["KernelD"],
case["KernelH"],
case["KernelW"],
case["InputD"],
case["InputH"],
case["InputW"],
case["StrideD"],
case["StrideH"],
case["StrideW"],
)
seen_2d = set()
unique_2d = []
for case in test_cases_2d:
@@ -320,7 +447,7 @@ def main():
if key not in seen_2d:
seen_2d.add(key)
unique_2d.append(case)
seen_3d = set()
unique_3d = []
for case in test_cases_3d:
@@ -328,11 +455,13 @@ def main():
if key not in seen_3d:
seen_3d.add(key)
unique_3d.append(case)
print(f"After deduplication: {len(unique_2d)} 2D cases, {len(unique_3d)} 3D cases")
print(
f"After deduplication: {len(unique_2d)} 2D cases, {len(unique_3d)} 3D cases"
)
test_cases_2d = unique_2d
test_cases_3d = unique_3d
# Write output files
if args.output:
# Write mixed cases to single file
@@ -340,14 +469,36 @@ def main():
if all_cases:
print(f"Writing {len(all_cases)} total cases to {args.output}")
# Use 2D headers for mixed file, extend as needed
mixed_headers = ['NDim', 'Groups', 'BatchSize', 'OutChannels', 'InChannels',
'KernelH', 'KernelW', 'InputH', 'InputW', 'OutputH', 'OutputW',
'StrideH', 'StrideW', 'DilationH', 'DilationW',
'LeftPadH', 'LeftPadW', 'RightPadH', 'RightPadW', 'TestName']
with open(args.output, 'w', newline='') as csvfile:
csvfile.write(f"# Mixed 2D/3D Convolution Test Cases from MIOpen Commands\n")
writer = csv.DictWriter(csvfile, fieldnames=mixed_headers, extrasaction='ignore')
mixed_headers = [
"NDim",
"Groups",
"BatchSize",
"OutChannels",
"InChannels",
"KernelH",
"KernelW",
"InputH",
"InputW",
"OutputH",
"OutputW",
"StrideH",
"StrideW",
"DilationH",
"DilationW",
"LeftPadH",
"LeftPadW",
"RightPadH",
"RightPadW",
"TestName",
]
with open(args.output, "w", newline="") as csvfile:
csvfile.write(
"# Mixed 2D/3D Convolution Test Cases from MIOpen Commands\n"
)
writer = csv.DictWriter(
csvfile, fieldnames=mixed_headers, extrasaction="ignore"
)
writer.writeheader()
for case in all_cases:
writer.writerow(case)
@@ -355,12 +506,13 @@ def main():
# Write separate files for 2D and 3D
if test_cases_2d:
write_csv_cases(test_cases_2d, args.output_2d, 2)
if test_cases_3d:
write_csv_cases(test_cases_3d, args.output_3d, 3)
print("Conversion completed!")
return 0
if __name__ == "__main__":
exit(main())

183
test_data/run_model_with_miopen.py
View File

@@ -7,13 +7,12 @@ PyTorch Model Runner with MIOpen Command Logging using torchvision models
Usage:
MIOPEN_ENABLE_LOGGING_CMD=1 python3 run_model_with_miopen.py --model resnet18 2> miopen_commands.txt
Available 2D models: alexnet, vgg11, vgg16, resnet18, resnet50, mobilenet_v2, etc.
Available 3D models: r3d_18, mc3_18, r2plus1d_18
"""
import torch
import torch.nn as nn
import torchvision.models as models
import torchvision.models.video as video_models
import argparse
@@ -21,94 +20,145 @@ import os
# Define available models
MODELS_2D = [
'alexnet', 'vgg11', 'vgg11_bn', 'vgg13', 'vgg13_bn', 'vgg16', 'vgg16_bn', 'vgg19', 'vgg19_bn',
'resnet18', 'resnet34', 'resnet50', 'resnet101', 'resnet152',
'resnext50_32x4d', 'resnext101_32x8d', 'resnext101_64x4d',
'wide_resnet50_2', 'wide_resnet101_2',
'densenet121', 'densenet161', 'densenet169', 'densenet201',
'inception_v3', 'googlenet',
'shufflenet_v2_x0_5', 'shufflenet_v2_x1_0', 'shufflenet_v2_x1_5', 'shufflenet_v2_x2_0',
'mobilenet_v2', 'mobilenet_v3_large', 'mobilenet_v3_small',
'mnasnet0_5', 'mnasnet0_75', 'mnasnet1_0', 'mnasnet1_3',
'squeezenet1_0', 'squeezenet1_1'
"alexnet",
"vgg11",
"vgg11_bn",
"vgg13",
"vgg13_bn",
"vgg16",
"vgg16_bn",
"vgg19",
"vgg19_bn",
"resnet18",
"resnet34",
"resnet50",
"resnet101",
"resnet152",
"resnext50_32x4d",
"resnext101_32x8d",
"resnext101_64x4d",
"wide_resnet50_2",
"wide_resnet101_2",
"densenet121",
"densenet161",
"densenet169",
"densenet201",
"inception_v3",
"googlenet",
"shufflenet_v2_x0_5",
"shufflenet_v2_x1_0",
"shufflenet_v2_x1_5",
"shufflenet_v2_x2_0",
"mobilenet_v2",
"mobilenet_v3_large",
"mobilenet_v3_small",
"mnasnet0_5",
"mnasnet0_75",
"mnasnet1_0",
"mnasnet1_3",
"squeezenet1_0",
"squeezenet1_1",
]
MODELS_3D = [
'r3d_18', 'mc3_18', 'r2plus1d_18'
]
MODELS_3D = ["r3d_18", "mc3_18", "r2plus1d_18"]
ALL_MODELS = MODELS_2D + MODELS_3D
def main():
parser = argparse.ArgumentParser(description='PyTorch Model Runner with MIOpen Command Logging')
parser = argparse.ArgumentParser(
description="PyTorch Model Runner with MIOpen Command Logging"
)
# Model selection
parser.add_argument('--model', choices=ALL_MODELS, default='resnet18',
help='Model to run')
parser.add_argument(
"--model", choices=ALL_MODELS, default="resnet18", help="Model to run"
)
# Input tensor dimensions
parser.add_argument('--batch-size', type=int, default=4,
help='Batch size')
parser.add_argument('--channels', type=int, default=3,
help='Input channels (e.g., 3 for RGB, 1 for grayscale)')
parser.add_argument('--height', type=int, default=224,
help='Input height')
parser.add_argument('--width', type=int, default=224,
help='Input width')
parser.add_argument('--input-size', type=int,
help='Input size (sets both height and width to same value)')
parser.add_argument('--temporal-size', type=int, default=16,
help='Temporal dimension for 3D models')
parser.add_argument("--batch-size", type=int, default=4, help="Batch size")
parser.add_argument(
"--channels",
type=int,
default=3,
help="Input channels (e.g., 3 for RGB, 1 for grayscale)",
)
parser.add_argument("--height", type=int, default=224, help="Input height")
parser.add_argument("--width", type=int, default=224, help="Input width")
parser.add_argument(
"--input-size",
type=int,
help="Input size (sets both height and width to same value)",
)
parser.add_argument(
"--temporal-size", type=int, default=16, help="Temporal dimension for 3D models"
)
# Device and precision
parser.add_argument('--device', choices=['cuda', 'cpu', 'auto'], default='auto',
help='Device to run on')
parser.add_argument('--precision', choices=['fp32', 'fp16', 'bf16'], default='fp32',
help='Floating point precision')
parser.add_argument(
"--device",
choices=["cuda", "cpu", "auto"],
default="auto",
help="Device to run on",
)
parser.add_argument(
"--precision",
choices=["fp32", "fp16", "bf16"],
default="fp32",
help="Floating point precision",
)
# Output control
parser.add_argument('--quiet', action='store_true',
help='Suppress output except errors')
parser.add_argument('--verbose', action='store_true',
help='Verbose output')
parser.add_argument(
"--quiet", action="store_true", help="Suppress output except errors"
)
parser.add_argument("--verbose", action="store_true", help="Verbose output")
args = parser.parse_args()
# Handle input-size override
if args.input_size:
args.height = args.input_size
args.width = args.input_size
# Check MIOpen logging
if not os.environ.get('MIOPEN_ENABLE_LOGGING_CMD') and not args.quiet:
if not os.environ.get("MIOPEN_ENABLE_LOGGING_CMD") and not args.quiet:
print("WARNING: Set MIOPEN_ENABLE_LOGGING_CMD=1 to capture commands")
# Device selection
if args.device == 'auto':
device = torch.device('cuda' if torch.cuda.is_available() else 'cpu')
if args.device == "auto":
device = torch.device("cuda" if torch.cuda.is_available() else "cpu")
else:
device = torch.device(args.device)
# Check if actually running on GPU
if device.type == 'cpu':
if device.type == "cpu":
import sys
print(f"WARNING: Running on CPU, MIOpen commands will not be generated!", file=sys.stderr)
print(
"WARNING: Running on CPU, MIOpen commands will not be generated!",
file=sys.stderr,
)
print(f"CUDA/ROCm available: {torch.cuda.is_available()}", file=sys.stderr)
if torch.cuda.is_available():
print(f"GPU device count: {torch.cuda.device_count()}", file=sys.stderr)
print(f"GPU name: {torch.cuda.get_device_name(0) if torch.cuda.device_count() > 0 else 'N/A'}", file=sys.stderr)
print(
f"GPU name: {torch.cuda.get_device_name(0) if torch.cuda.device_count() > 0 else 'N/A'}",
file=sys.stderr,
)
# Continue anyway for testing purposes
if not args.quiet:
print(f"Using device: {device}")
# Create model using torchvision
if args.model in MODELS_3D:
# 3D Video models
model = getattr(video_models, args.model)(weights=None)
# 3D input: (batch, channels, temporal, height, width)
input_tensor = torch.randn(args.batch_size, args.channels, args.temporal_size, args.height, args.width)
input_tensor = torch.randn(
args.batch_size, args.channels, args.temporal_size, args.height, args.width
)
if not args.quiet:
print(f"3D model: {args.model}")
print(f"Input shape: {input_tensor.shape} (B, C, T, H, W)")
@@ -116,34 +166,37 @@ def main():
# 2D Image models
model = getattr(models, args.model)(weights=None)
# 2D input: (batch, channels, height, width)
input_tensor = torch.randn(args.batch_size, args.channels, args.height, args.width)
input_tensor = torch.randn(
args.batch_size, args.channels, args.height, args.width
)
if not args.quiet:
print(f"2D model: {args.model}")
print(f"Input shape: {input_tensor.shape} (B, C, H, W)")
# Set precision
if args.precision == 'fp16':
if args.precision == "fp16":
model = model.half()
input_tensor = input_tensor.half()
elif args.precision == 'bf16':
elif args.precision == "bf16":
model = model.bfloat16()
input_tensor = input_tensor.bfloat16()
model = model.to(device)
input_tensor = input_tensor.to(device)
if not args.quiet:
print(f"Running {args.model} model...")
# Run inference
model.eval()
with torch.no_grad():
output = model(input_tensor)
if not args.quiet:
print(f"Output shape: {output.shape}")
if not args.quiet:
print("Done! MIOpen commands logged to stderr")
if __name__ == "__main__":
main()