[MIOpen] Add JSON performance logs for MIOpen convolution driver commands by jdcampbe · Pull Request #5736 · ROCm/rocm-libraries

jdcampbe · 2026-03-23T23:41:10Z

Adds a performance logging system to MIOpen for tracking kernel names and execution times. This enables debugging, performance analysis, and understanding which kernels execute under different configurations. Reduces errors caused by parsing the logs and enables tracking kernel executions and times.

Motivation

MIOpen needs an option to gather granular performance data for heuristics, evaluation, and debugging. MIOpen current logging format is human readable but difficult to parse due to the ad hoc nature of the logs. The new performance logs introduce JSON objects that can be found and parse more easily.

Technical Details

Introduces MIOPEN_PERFORMANCE_LOGS environment variable with 5 logging levels (0-4):

Level 0: No logging (default)
Level 1: Logs the Chosen/Executed solution
Level 2: All loggable kernel information is include for the Chosen/Executed solution
Level 3: Logs the Chosen/Executed, and Tuning/Search solutions
Level 4: All loggable kernel information is include for the Chosen/Executed, and Tuning/Search solutions

Key Changes:

New kernel_tuning_mode.hpp with logging infrastructure and JSON output support
Tuning timing tracked through conv_driver.cpp
Phase tracking (execution, validation, tuning) integrated into solver finders and generic search
Solution-level timing aggregation with statistics (mean, std, min, max)
Thread-safe implementation using thread-local storage
Comprehensive documentation in driver/README.md

Test Plan

Verified convolution logging through data gathering. All logging levels validated. JSON output verified. Existing logs are unimpacted (as long as the MIOPEN_PERFORMANCE_LOGS environment variable is set to 0 or left at the default.

Submission Checklist

Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.

- Add format for performance logs - Convert to more focused name "MIOPEN_PERFORMANCE_LOGS" - Update licenses

- Fix typo for bytes writter to bytes written - Fix build errors.

- Increase verbose logging for transformation kernels.

- Add calculations - Fix variable naming

- Increment iterations properly - Log solution name in result with DB key - Fix logging for WRW BWD and FWD

- log env variables and find time in forward.

- Add find_time logs for Backwards Data - Add find_time logs for Backwards Weights

- Update to find_time_conv - Fix missing references to perf log

- Update the logs to solution->performance_configs->kernels - Add missing transformation kernels - Move timing logic to the performance config

- Update phase tuning to be more accurate - Refactor IsKernelLogging to reduce redundancy.

tkoermen-amd

Okay, approving this pending CI. Thanks so much for resolving all the comments, I appreciate it.

- Update perf DB key to match the used DB key

…pen-log-kernel-names

Co-Authored-By: Claude Opus 4 <noreply@anthropic.com>

Covers env-var gating across log levels 0-4, ScopedKernelPhase RAII, JSON emission to stderr, and accumulator no-op behavior when logging is disabled. CPU-only, registered as CPU_KernelTuningModeLogs_NONE. Co-Authored-By: Claude Opus 4 <noreply@anthropic.com>

JIRA ID: ALMIOPEN-2312 ## Motivation `projects/miopen/CHANGELOG.md`'s "MIOpen 3.5.2 for ROCm 7.14.0" section only had 4 entries, but 68 commits touched `projects/miopen` between the `therock-7.13` tag and `release/therock-7.14`. Auditing that commit range (via `git log --oneline therock-7.13..remotes/origin/release/therock-7.14 -- projects/miopen`) found 18 additional user-facing changes — new gfx950/gfx1250 architecture support, SystemDB refreshes, AI-heuristic improvements, several correctness/crash bug fixes, and OpenCL solver removals — that were never documented. This PR backfills those entries so the 7.14.0 changelog accurately reflects what shipped. ## Technical Details Added 24 lines to the existing "MIOpen 3.5.2 for ROCm 7.14.0" section of `projects/miopen/CHANGELOG.md`, merging new bullets into the existing `Added`/`Changed`/`Resolved Issues` sub-sections (no new entries were duplicated or reworded) and adding a new `Removed` sub-section for solver removals. No existing content was altered. Each new bullet was traced to a specific commit + PR: | Entry | Commit | PR | |---|---|---| | Added gfx950 (MI350X/MI355X) SystemDB | `60d8e1e5175` | #8081 | | Added initial gfx1250 architecture support | `d3f057b2f22` | #7587 | | Added JSON-formatted performance logs (`MIOPEN_PERFORMANCE_LOGS`) | `ffbde871260` | #5736 | | Refreshed gfx942 SystemDB | `faf754a9ef6` | #8174 | | Refined 2D AI heuristics (Tunanet/KTN) for gfx942 & gfx950 | `c24ad415884` | #7109 | | Fixed thread-safety issue in AI-heuristic model caches | `cd957402309` | #9149 (cherry-pick of #9131) | | Fixed CK grouped-conv solvers not always registered in host | `c971d54d7ae` | #9105 | | Fixed legacy CK buffer resource descriptor (gfx1250) | `1dee011b8b0` | #8758 | | Removed stale gfx942 SystemDB naive-solver entries | `f42668f6316` | #7028 | | Fixed softmax NaN when `beta=0` | `fb8157c8512` | #8426 | | Fixed BN backward spatial variant 1 vectorized memory access bug | `1351a7e8b10` | #6765 | | Fixed use-after-free of `kernel_id` in `InitInvokerFactoryNCHW` | `394ebffb50d` | #7922 | | Fixed SoftMarginLoss forward reduce-kernel kernarg corruption | `1e65a31570f` | #8065, #8135 | | Fixed build with `MIOPEN_ENABLE_AI_KERNEL_TUNING=OFF` | `e6288c3ca61` | #6800 | | Fixed install RPATH for MIOpenDriver / CK backend libs | `ef8a4cf2f93` | #7485 | | Fixed dangling lambda causing log-level 6 segfault | `190f9791737` | #6953 | | Enabled Windows long-path support for `addkernels` | `1f9d69724cf` | #7387 | | Removed OCL solver `ConvOclDirectFwd1x1` | `126c720cadd` | #6491 | | Removed OCL solver `conv_ocl_dir2Dfwdgen` | `2d499ddf709` | #6633 | | Removed OCL solvers `ConvOclBwdWrW1x1`, `ConvOclBwdWrW2` | `b5d43a33dc6` | #6859 | | Removed OCL solvers `OpTensorLeadingOnes`, `Op5dTensorGeneric` | `8642c83b7de` | #7347 | The remaining ~40 commits in the 68-commit range were deliberately excluded (tests/CI/build/docs/internal refactors/1:1 OCL→HIP swaps with no behavior change, per each commit's own stated scope) and are not represented here. ## Risk Assessment **Risk Level: 🟢 Low** This is a documentation-only change to a single Markdown file. No code, build configuration, tests, or public API surface is touched. ### Impacted Components - `projects/miopen/CHANGELOG.md` only. ### Potential Side Effects None expected. This does not change any runtime behavior; it only documents changes that already shipped on `release/therock-7.14`. ### Mitigation Steps None required beyond normal review — a reviewer familiar with the 7.14 release can spot-check a few entries against the linked commit hashes/PR numbers for accuracy. ## Test Plan Not applicable (no code changes). Verification was done by cross-referencing every added bullet against a specific commit hash and PR number pulled from `git log`, and by diffing the edited section against `git show remotes/origin/release/therock-7.14:projects/miopen/CHANGELOG.md` to confirm no existing entry was duplicated or altered. ## Test Result All 21 new bullets map 1:1 to a real commit/PR in the `therock-7.13..release/therock-7.14` range for `projects/miopen`. No pre-existing changelog text was modified — `git diff` shows a pure addition (24 insertions, 0 deletions). Two borderline commits (`a67834413e8` "Standalone build", `7c7cd5774a9` "enable warnings about deprecated API/features") were deliberately left out pending author judgment on whether they warrant an entry.

jdcampbe added 30 commits February 4, 2026 01:33

Log kernel names when environment variables present

84ca11a

JSON Logs

5d18a2f

- Add format for performance logs - Convert to more focused name "MIOPEN_PERFORMANCE_LOGS" - Update licenses

Fix Typos

9cd0d40

- Fix typo for bytes writter to bytes written - Fix build errors.

Fix Verbose Logging

2677ef7

- Increase verbose logging for transformation kernels.

Compiler Fixes

3bd39f3

Add tuning flag to evaluate invokers

349f14b

Add Verification to Tuning

f90af96

Update variable naming.

00aee5e

- Add calculations - Fix variable naming

Add missing logging for WRW BWD

ad0cb4f

- Increment iterations properly - Log solution name in result with DB key - Fix logging for WRW BWD and FWD

Log find time in forward

a59f311

- log env variables and find time in forward.

Find time

bbe03ab

- Add find_time logs for Backwards Data - Add find_time logs for Backwards Weights

Update variable name

18a5248

- Update to find_time_conv - Fix missing references to perf log

Add DB Key to performance logs

96a210a

Update performance logs solution to include solution id

3afa61c

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

5dbaea8

Fix function arguments mismatch

1822993

Fix scoping error

ac0b54c

Update db key to match system db

64ed4e1

Log solution if solution id changes

188d616

Set time on if performance logs are active.

4ea7785

Set timing in different location so it is not overwritten

b3a6a1c

Add Performance Configs

42875f7

- Update the logs to solution->performance_configs->kernels - Add missing transformation kernels - Move timing logic to the performance config

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

eab71b7

Fix compile issues and missing solutions.

ecef3f6

Add more complex phases to performance logs

2fd2748

Modify Phase Tuning

7acda22

- Update phase tuning to be more accurate - Refactor IsKernelLogging to reduce redundancy.

Fix compiler errors

031f950

Fix kernels not showing up

1b71d3c

Remove duplicate code

8b471de

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

191e5c1

jdcampbe added 2 commits April 1, 2026 14:44

Reorder invoker times gathering to avoid the changes to the sample list.

d535ec9

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

6a0867e

tkoermen-amd approved these changes Apr 1, 2026

View reviewed changes

jdcampbe and others added 22 commits April 2, 2026 15:35

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

304a220

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

d7916ba

Update Perf DB Key

16a17c9

- Update perf DB key to match the used DB key

Merge remote-tracking branch 'origin/develop' into users/jascampb/mio…

3dc5cd0

…pen-log-kernel-names

Match kernel name.

d90e22f

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

1dc674f

Fix dereferencing error

70096de

Apply clang-format to conv_driver.hpp

f02f139

Co-Authored-By: Claude Opus 4 <noreply@anthropic.com>

Removed unused variables from json call site.

00a8aba

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

b0e2b8f

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

7260944

Fix tidy errors on kernel tuning mode

0373345

Fix clang tidy.

b2cedd2

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

7fa1fb0

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

6a2b7e0

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

34decf0

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

0264aa3

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

441fd7a

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

ff01136

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

c700688

Merge branch 'develop' into users/jascampb/miopen-log-kernel-names

9db59a1

jdcampbe merged commit ffbde87 into develop May 26, 2026
40 checks passed

jdcampbe deleted the users/jascampb/miopen-log-kernel-names branch May 26, 2026 15:38

JonathanLichtnerAMD mentioned this pull request Jul 15, 2026

docs(miopen): backfill missing 7.14.0 changelog entries #9433

Merged

Navigation Menu

Repository navigation

[MIOpen] Add JSON performance logs for MIOpen convolution driver commands - #5736

[MIOpen] Add JSON performance logs for MIOpen convolution driver commands#5736
jdcampbe merged 62 commits into
developfrom
users/jascampb/miopen-log-kernel-names

jdcampbe commented Mar 23, 2026

Uh oh!

tkoermen-amd left a comment

Uh oh!

Uh oh!

Reviewers

Uh oh!

Assignees

Labels

Projects

Milestone

Development

Uh oh!

4 participants

Conversation

jdcampbe commented Mar 23, 2026

Motivation

Technical Details

Test Plan

Submission Checklist

Uh oh!

tkoermen-amd left a comment

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Reviewers

Uh oh!

Assignees

Labels

Projects

Milestone

Development

Uh oh!

4 participants