Files
derek.holloway 7502bf64a4
CodeQL / Analyze (actions) (push) Canceled after 0s
CodeQL / Analyze (python) (push) Canceled after 0s
pre-commit / pre-commit (push) Canceled after 0s
Multi-Arch CI / setup (push) Canceled after 0s
Publish build_manylinux_x86_64 images / publish_build_manylinux_x86_64 (push) Canceled after 0s
Setup (Multi-Arch) / setup (push) Canceled after 0s
Multi-Arch CI / Linux::${{ fromJSON(needs.setup.outputs.linux_build_config || '{}').build_variant_label || 'skip' }} (push) Canceled after 0s
Publish TheRock Docker image / build-and-push-image (push) Canceled after 0s
Publish no_rocm_image_ubi10 image / publish_no_rocm_image_ubi10 (push) Canceled after 0s
Multi-Arch CI / Windows::${{ fromJSON(needs.setup.outputs.windows_build_config || '{}').build_variant_label || 'skip' }} (push) Canceled after 0s
Publish TheRock Docker image / build-and-push-image (push) Canceled after 0s
Multi-Arch CI / Manifest Diff (push) Canceled after 0s
TheRock Manifest Diff Report / Generate Manifest Diff Report (push) Canceled after 0s
Multi-Arch CI / CI Summary (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04 images / publish_no_rocm_image_ubuntu24_04 (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04_ocl_rt images / publish_no_rocm_image_ubuntu24_04_ocl_rt (push) Canceled after 0s
Publish TheRock Docker image / build-and-push-image (push) Canceled after 0s
Publish TheRock Docker image / build-and-push-image (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04 images / publish_no_rocm_image_ubuntu24_04_rocgdb (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04_rocgdb images / publish_no_rocm_image_ubuntu24_04_rocgdb (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04 images / publish_no_rocm_image_ubuntu24_04_ocl_rt (push) Canceled after 0s
Publish TheRock Docker image / build-and-push-image (push) Canceled after 0s
Unit Tests / Unit Tests :: ubuntu-24.04 (push) Canceled after 0s
Unit Tests / Unit Tests :: windows-2022 (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04 images / publish_no_rocm_image_ubuntu24_04_media (push) Canceled after 0s
Unit Tests / Unit Tests Summary (push) Canceled after 0s
Publish no_rocm_image_ubuntu24_04 images / publish_no_rocm_image_ubuntu24_04_openmpi (push) Canceled after 0s
Flatten subprojects into a single base monorepo
2026-09-02 16:34:47 -07:00

566 lines
17 KiB
Python

# Copyright (c) Advanced Micro Devices, Inc.
# SPDX-License-Identifier: MIT
import json
import sqlite3
from pathlib import Path
import common
import pandas as pd
from utils.rocpd_data import (
COUNTERS_COLLECTION_QUERY,
MARKER_API_TRACE_QUERY,
convert_dbs_to_csv,
)
from utils.utils_analysis import (
build_call_trees_with_kernel_ids,
process_ml_api_trace_output,
write_ml_api_trace_consolidated_csv,
)
from utils.utils_profile import _augment_marker_csv, _parse_function_backend
GUID = "abc-1234-def"
MARKER_ROWS = [
(
"roctx",
"nn.Module.Linear.forward:#1@test.py:10",
100,
200,
1000,
GUID,
1000,
2000,
),
(
"roctx",
"nn.Module.Linear.forward:#2@test.py:10",
100,
200,
1001,
GUID,
3000,
4000,
),
("roctx", "torch.mm:#1@test.py:15", 100, 200, 1002, GUID, 5000, 6000),
]
COUNTER_ROWS = [
(
0,
GUID,
1000,
0,
100,
64,
256,
0,
0,
32,
0,
16,
"kernel_gemm",
1100,
1900,
0,
"SQ_WAVES",
42,
),
(
0,
GUID,
1001,
1,
100,
64,
256,
0,
0,
32,
0,
16,
"kernel_gemm",
3100,
3900,
0,
"SQ_WAVES",
50,
),
(
0,
GUID,
1002,
2,
100,
64,
256,
0,
0,
32,
0,
16,
"kernel_mm",
5100,
5900,
0,
"SQ_WAVES",
30,
),
]
# ---- SQL query constants reference stack_id ----
def test_counters_query_uses_stack_id():
"""Test that the counters query uses stack_id as Correlation_Id."""
assert "stack_id as Correlation_Id" in COUNTERS_COLLECTION_QUERY
query_lower = COUNTERS_COLLECTION_QUERY.lower()
assert "correlation_id as " not in query_lower
assert "\n correlation_id" not in query_lower
def test_marker_query_uses_stack_id():
"""Test that the marker query uses stack_id as Correlation_Id."""
assert "stack_id AS Correlation_Id" in MARKER_API_TRACE_QUERY
query_lower = MARKER_API_TRACE_QUERY.lower()
assert "correlation_id as " not in query_lower
assert "\n correlation_id" not in query_lower
# ---- Test 2: convert_dbs_to_csv populates Correlation_Id from stack_id ----
def create_rocpd_test_db(workload_dir):
"""
Build a minimal rocpd-style SQLite database with counters_collection
and regions tables whose schemas match the production queries.
"""
db_path = str(Path(workload_dir) / "test.db")
conn = sqlite3.connect(db_path)
conn.execute(
"""CREATE TABLE counters_collection (
agent_id INTEGER, guid TEXT, stack_id INTEGER, dispatch_id INTEGER,
pid INTEGER, grid_size INTEGER, workgroup_size INTEGER,
lds_block_size INTEGER, scratch_size INTEGER, vgpr_count INTEGER,
accum_vgpr_count INTEGER, sgpr_count INTEGER, kernel_name TEXT,
start INTEGER, end INTEGER, kernel_id INTEGER,
counter_name TEXT, value REAL
)"""
)
conn.executemany(
"INSERT INTO counters_collection VALUES (?,?,?,?,?,?,?,?,?,?,?,?,?,?,?,?,?,?)",
COUNTER_ROWS,
)
conn.execute(
"""CREATE TABLE regions (
category TEXT, extdata TEXT, pid INTEGER, tid INTEGER,
stack_id INTEGER, guid TEXT, start INTEGER, end INTEGER
)"""
)
region_rows = [
(cat, json.dumps({"message": func}), pid, tid, sid, guid, s, e)
for cat, func, pid, tid, sid, guid, s, e in MARKER_ROWS
]
conn.executemany(
"INSERT INTO regions VALUES (?,?,?,?,?,?,?,?)",
region_rows,
)
conn.commit()
conn.close()
return db_path
def test_counter_csv_has_correlation_id_from_stack_id():
"""Test that the counter CSV has correlation_id from stack_id."""
workload_dir = common.get_output_dir()
Path(workload_dir).mkdir(parents=True, exist_ok=True)
counter_csv = str(Path(workload_dir) / "counter_collection.csv")
marker_csv = str(Path(workload_dir) / "marker_api_trace.csv")
db_path = create_rocpd_test_db(workload_dir)
convert_dbs_to_csv([db_path], counter_csv, marker_csv)
df = pd.read_csv(counter_csv)
assert "Correlation_Id" in df.columns
expected_ids = [row[2] for row in COUNTER_ROWS]
assert list(df["Correlation_Id"]) == expected_ids
common.clean_output_dir(True, workload_dir)
def test_marker_csv_has_correlation_id_from_stack_id():
"""Test that the marker CSV has correlation_id from stack_id."""
workload_dir = common.get_output_dir()
Path(workload_dir).mkdir(parents=True, exist_ok=True)
counter_csv = str(Path(workload_dir) / "counter_collection.csv")
marker_csv = str(Path(workload_dir) / "marker_api_trace.csv")
db_path = create_rocpd_test_db(workload_dir)
convert_dbs_to_csv([db_path], counter_csv, marker_csv)
df = pd.read_csv(marker_csv)
assert "Correlation_Id" in df.columns
expected_ids = sorted(row[4] for row in MARKER_ROWS)
assert sorted(df["Correlation_Id"].tolist()) == expected_ids
common.clean_output_dir(True, workload_dir)
# ---- process_ml_api_trace_output parity for rocpd vs csv layouts ----
MARKER_COLUMNS_ROCPD = [
"Domain",
"Function",
"Process_Id",
"Thread_Id",
"Correlation_Id",
"GUID",
"Start_Timestamp",
"End_Timestamp",
]
COUNTER_COLUMNS_ROCPD = [
"GPU_ID",
"GUID",
"Correlation_Id",
"Dispatch_ID",
"PID",
"Grid_Size",
"Workgroup_Size",
"LDS_Per_Workgroup",
"Scratch_Per_Workitem",
"Arch_VGPR",
"Accum_VGPR",
"SGPR",
"Kernel_Name",
"Start_Timestamp",
"End_Timestamp",
"Kernel_ID",
"Counter_Name",
"Counter_Value",
]
MARKER_COLUMNS_CSV = [
"Domain",
"Function",
"Process_Id",
"Thread_Id",
"Correlation_Id",
"Start_Timestamp",
"End_Timestamp",
]
COUNTER_COLUMNS_CSV = [
"Correlation_Id",
"Kernel_Name",
"Counter_Name",
"Counter_Value",
"Start_Timestamp",
"End_Timestamp",
]
def build_marker_df(include_guid):
"""Build a dataframe from the marker rows."""
data = {
"Domain": [r[0] for r in MARKER_ROWS],
"Function": [r[1] for r in MARKER_ROWS],
"Process_Id": [r[2] for r in MARKER_ROWS],
"Thread_Id": [r[3] for r in MARKER_ROWS],
"Correlation_Id": [r[4] for r in MARKER_ROWS],
"Start_Timestamp": [r[6] for r in MARKER_ROWS],
"End_Timestamp": [r[7] for r in MARKER_ROWS],
}
if include_guid:
data["GUID"] = [r[5] for r in MARKER_ROWS]
return pd.DataFrame(data)
def build_counter_df(include_guid):
"""Build a dataframe from the counter rows."""
data = {
"Correlation_Id": [r[2] for r in COUNTER_ROWS],
"Kernel_Name": [r[12] for r in COUNTER_ROWS],
"Counter_Name": [r[16] for r in COUNTER_ROWS],
"Counter_Value": [r[17] for r in COUNTER_ROWS],
"Start_Timestamp": [r[13] for r in COUNTER_ROWS],
"End_Timestamp": [r[14] for r in COUNTER_ROWS],
}
if include_guid:
data["GUID"] = [r[1] for r in COUNTER_ROWS]
data["GPU_ID"] = [r[0] for r in COUNTER_ROWS]
data["Dispatch_ID"] = [r[3] for r in COUNTER_ROWS]
data["PID"] = [r[4] for r in COUNTER_ROWS]
data["Grid_Size"] = [r[5] for r in COUNTER_ROWS]
data["Workgroup_Size"] = [r[6] for r in COUNTER_ROWS]
data["LDS_Per_Workgroup"] = [r[7] for r in COUNTER_ROWS]
data["Scratch_Per_Workitem"] = [r[8] for r in COUNTER_ROWS]
data["Arch_VGPR"] = [r[9] for r in COUNTER_ROWS]
data["Accum_VGPR"] = [r[10] for r in COUNTER_ROWS]
data["SGPR"] = [r[11] for r in COUNTER_ROWS]
data["Kernel_ID"] = [r[15] for r in COUNTER_ROWS]
return pd.DataFrame(data)
def write_rocpd_layout(workload_dir, fbase="run0"):
"""Write marker/counter CSVs at workload root (rocpd layout)."""
marker_df = build_marker_df(include_guid=True)
counter_df = build_counter_df(include_guid=True)
marker_path = Path(workload_dir) / f"ml_api_trace_{fbase}_marker_api_trace.csv"
counter_path = Path(workload_dir) / f"ml_api_trace_{fbase}_counter_collection.csv"
marker_df.to_csv(marker_path, index=False)
counter_df.to_csv(counter_path, index=False)
def write_csv_layout(workload_dir, fbase="run0", pid="12345"):
"""Write marker/counter CSVs in a subdirectory (csv layout)."""
subdir = Path(workload_dir) / fbase
subdir.mkdir(parents=True, exist_ok=True)
marker_df = build_marker_df(include_guid=False)
counter_df = build_counter_df(include_guid=False)
marker_path = subdir / f"ml_api_trace_{pid}_marker_api_trace.csv"
counter_path = subdir / f"ml_api_trace_{pid}_counter_collection.csv"
marker_df.to_csv(marker_path, index=False)
counter_df.to_csv(counter_path, index=False)
def read_ml_api_trace_csvs(ml_api_trace_dir):
"""Return a dict mapping filename -> sorted DataFrame for comparison."""
result = {}
for csv_file in sorted(Path(ml_api_trace_dir).glob("*.csv")):
df = pd.read_csv(csv_file)
df = df.sort_values(by=list(df.columns)).reset_index(drop=True)
result[csv_file.name] = df
return result
def build_kernel_top_df():
"""Build a kernel top stats DataFrame matching the test kernel names."""
return pd.DataFrame({
"Kernel_Name": sorted({r[12] for r in COUNTER_ROWS}),
})
def test_ml_api_trace_output_same_for_rocpd_and_csv():
"""Test that the ML API trace output is the same for rocpd and csv files."""
rocpd_dir = common.get_output_dir(suffix="_rocpd")
csv_dir = common.get_output_dir(suffix="_csv")
Path(rocpd_dir).mkdir(parents=True, exist_ok=True)
Path(csv_dir).mkdir(parents=True, exist_ok=True)
write_rocpd_layout(rocpd_dir)
write_csv_layout(csv_dir)
kernel_top_df = build_kernel_top_df()
rocpd_output = process_ml_api_trace_output(rocpd_dir)
csv_output = process_ml_api_trace_output(csv_dir)
assert rocpd_output is not None
assert csv_output is not None
rocpd_df, rocpd_trace_path = rocpd_output
csv_df, csv_trace_path = csv_output
write_ml_api_trace_consolidated_csv(rocpd_df, rocpd_trace_path)
write_ml_api_trace_consolidated_csv(csv_df, csv_trace_path)
rocpd_trees = build_call_trees_with_kernel_ids(rocpd_df, kernel_top_df)
csv_trees = build_call_trees_with_kernel_ids(csv_df, kernel_top_df)
for trees in (rocpd_trees, csv_trees):
assert "test.py:10" in trees
assert "test.py:15" in trees
linear_root = trees["test.py:10"]
assert linear_root.kernel_launches == 2
assert "nn.Module.Linear.forward" in linear_root.children
linear_node = linear_root.children["nn.Module.Linear.forward"]
assert "kernel_gemm" in linear_node.kernels
assert linear_node.kernels["kernel_gemm"].launches == 2
mm_root = trees["test.py:15"]
assert mm_root.kernel_launches == 1
assert "torch.mm" in mm_root.children
mm_node = mm_root.children["torch.mm"]
assert "kernel_mm" in mm_node.kernels
assert mm_node.kernels["kernel_mm"].launches == 1
rocpd_results = read_ml_api_trace_csvs(Path(rocpd_dir) / "ml_api_trace")
csv_results = read_ml_api_trace_csvs(Path(csv_dir) / "ml_api_trace")
assert rocpd_results.keys() == csv_results.keys(), (
f"ML API trace CSV files differ: rocpd={sorted(rocpd_results.keys())} "
f"csv={sorted(csv_results.keys())}"
)
for filename in rocpd_results:
pd.testing.assert_frame_equal(
rocpd_results[filename],
csv_results[filename],
check_dtype=False,
obj=filename,
)
common.clean_output_dir(True, rocpd_dir)
common.clean_output_dir(True, csv_dir)
# ---- Backend column unpacking in save_ml_api_trace_inputs ----
def test_parse_function_backend_untagged_is_unknown():
"""Untagged rows surface as Backend='unknown'."""
clean, backend = _parse_function_backend("torch.empty:#1@linear.py:109")
assert clean == "torch.empty:#1@linear.py:109"
assert backend == "unknown"
def test_parse_function_backend_tagged_torch_is_stripped():
"""Tagged single-frame markers expose backend and lose the suffix."""
clean, backend = _parse_function_backend(
"nn.Module.MyModel.forward:#1@train.py:42|torch"
)
assert clean == "nn.Module.MyModel.forward:#1@train.py:42"
assert backend == "torch"
def test_parse_function_backend_tagged_triton_leaf():
"""Row-level suffix attributes the entire wire to its producing backend."""
clean, backend = _parse_function_backend(
"torch.compile.fn/triton.CompiledKernel.foo:#1@a.py:1/#1@b.py:2|triton"
)
assert clean == ("torch.compile.fn/triton.CompiledKernel.foo:#1@a.py:1/#1@b.py:2")
assert backend == "triton"
def test_parse_function_backend_aten_leaf_is_unknown():
"""Untagged ATen leaf surfaces as Backend='unknown'."""
clean, backend = _parse_function_backend(
"nn.Module.X.forward/aten::add:#1@m.py:9/#1@aten:0"
)
assert clean == "nn.Module.X.forward/aten::add:#1@m.py:9/#1@aten:0"
assert backend == "unknown"
def test_parse_function_backend_edge_cases():
"""Bogus suffix, empty string, and None all fall back to 'unknown'."""
assert _parse_function_backend("op|bogus") == ("op|bogus", "unknown")
assert _parse_function_backend("") == ("", "unknown")
assert _parse_function_backend(None) == ("", "unknown")
def test_augment_marker_csv_untagged_row_warns(tmp_path, monkeypatch):
"""Untagged rows are tagged 'unknown' and emit a warning."""
from utils import utils_profile
src = tmp_path / "src_marker_api_trace.csv"
dst = tmp_path / "ml_api_trace_dst_marker_api_trace.csv"
pd.DataFrame({"Function": ["aten::sum"]}).to_csv(src, index=False)
warnings: list[tuple] = []
monkeypatch.setattr(utils_profile, "console_warning", lambda *a: warnings.append(a))
_augment_marker_csv(str(src), str(dst))
out_df = pd.read_csv(dst)
assert out_df["Function"].tolist() == ["aten::sum"]
assert out_df["Backend"].tolist() == ["unknown"]
assert warnings, "untagged rows must emit a warning"
assert any("unknown" in str(a) for a in warnings[0])
def test_augment_marker_csv_adds_backend_column(tmp_path):
"""End-to-end: tagged + untagged rows survive copy; Backend is populated."""
src = tmp_path / "src_marker_api_trace.csv"
dst = tmp_path / "ml_api_trace_dst_marker_api_trace.csv"
src_df = pd.DataFrame({
"Domain": ["MARKER_CORE_RANGE_API"] * 3,
"Function": [
"nn.Module.X.forward:#1@a.py:1|torch",
"triton.CompiledKernel.k:#1@b.py:2|triton",
"torch.empty:#1@c.py:3",
],
"Correlation_Id": [1, 2, 3],
"Start_Timestamp": [100, 200, 300],
"End_Timestamp": [150, 250, 350],
})
src_df.to_csv(src, index=False)
_augment_marker_csv(str(src), str(dst))
out_df = pd.read_csv(dst)
assert "Backend" in out_df.columns
assert out_df["Backend"].tolist() == ["torch", "triton", "unknown"]
assert out_df["Function"].tolist() == [
"nn.Module.X.forward:#1@a.py:1",
"triton.CompiledKernel.k:#1@b.py:2",
"torch.empty:#1@c.py:3",
]
for col in ("Domain", "Correlation_Id", "Start_Timestamp", "End_Timestamp"):
assert col in out_df.columns
def test_augment_marker_csv_handles_unknown_schema(tmp_path):
"""A CSV without a Function column copies verbatim instead of corrupting."""
src = tmp_path / "src.csv"
dst = tmp_path / "dst.csv"
src.write_text("Foo,Bar\n1,2\n3,4\n", encoding="utf-8")
_augment_marker_csv(str(src), str(dst))
assert dst.read_text(encoding="utf-8") == src.read_text(encoding="utf-8")
def test_process_ml_api_trace_output_defaults_backend_for_untagged(tmp_path):
"""Untagged rows default to Backend='torch' in the consolidated df."""
workload_dir = str(tmp_path)
write_rocpd_layout(workload_dir)
consolidated_df, _ = process_ml_api_trace_output(workload_dir)
assert "Backend" in consolidated_df.columns
assert (consolidated_df["Backend"] == "torch").all()
def test_process_ml_api_trace_output_preserves_per_row_backend(tmp_path):
"""A pre-stripped + tagged CSV (as produced by _augment_marker_csv)
surfaces the per-row Backend value into the consolidated dataframe.
"""
workload_dir = str(tmp_path)
write_rocpd_layout(workload_dir)
# Overwrite the fixture with what save_ml_api_trace_inputs would produce:
# Function has prefixes stripped, Backend carries per-row attribution.
marker_path = Path(workload_dir) / "ml_api_trace_run0_marker_api_trace.csv"
df = pd.read_csv(marker_path)
df["Backend"] = ["torch", "torch", "triton"]
df.to_csv(marker_path, index=False)
consolidated_df, _ = process_ml_api_trace_output(workload_dir)
assert "Backend" in consolidated_df.columns
backend_by_operator = dict(
zip(consolidated_df["Operator_Name"], consolidated_df["Backend"])
)
assert backend_by_operator.get("torch.mm") == "triton"
assert backend_by_operator.get("nn.Module.Linear.forward") == "torch"