gdb/amd-dbgapi-target: add trap-on-entry support - #219
Open
sebdar wants to merge 1 commit into
Open
Conversation
Contributor
Author
|
Corresponding dbgapi PR : ROCm/rocm-systems#8762 |
aktemur
reviewed
Jul 21, 2026
Collaborator
|
@sebdar Is there a ticket tracking this work? If so, it would be nice to add it to the PR/commit. |
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
July 22, 2026 14:43
30b0af2 to
3a8e4e5
Compare
Contributor
Author
|
Thanks for the review @aktemur . In the last push :
|
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
July 22, 2026 14:48
3a8e4e5 to
2506a8c
Compare
aktemur
approved these changes
Jul 24, 2026
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
2 times, most recently
from
July 27, 2026 19:05
d6849ff to
ceee9db
Compare
lancesix
requested changes
Jul 28, 2026
lancesix
left a comment
lancesix
left a comment
Collaborator
There was a problem hiding this comment.
Documentation is missing.
Other bits and bobs comments below.
| # Not every architecture supports trap-on-entry. Request the feature before | ||
| # the process starts, then let the process attach apply it and report through | ||
| # "show" whether it actually became effective. | ||
| set trap_on_entry_supported false |
Collaborator
There was a problem hiding this comment.
nit: I think I would make supports_trap_on_entry a function which can be called, rather than inlining it. But this is mostly a me thing.
| static void | ||
| set_process_trap_on_entry (amd_dbgapi_inferior_info &info) | ||
| { | ||
| #if AMD_DBGAPI_VERSION_MAJOR > 0 || AMD_DBGAPI_VERSION_MINOR >= 81 |
Collaborator
There was a problem hiding this comment.
Ideally, we might want to have a configure test checking if AMD_DBGAPI_WAVE_LAUNCH_TRAP_ON_ENTRY_ENABLED is defined for example, and set a AMD_DBGAPI_SUPPORTS_TRAP_ON_ENTRY. This would remove some ordering issues, as we have quite a few dbgapi things in the pipeline comming, ordering might not guarantee that this feature will be 0.81.
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
August 3, 2026 20:50
ceee9db to
539883b
Compare
Contributor
Author
|
In the last push :
|
lancesix
reviewed
Aug 3, 2026
lancesix
left a comment
lancesix
left a comment
Collaborator
There was a problem hiding this comment.
some minor comments, otherwise LGTM.
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
August 4, 2026 13:36
539883b to
655de73
Compare
Contributor
Author
|
In the last push :
|
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
August 10, 2026 18:44
655de73 to
eb00ed3
Compare
This commit introduces interface to the new 'trap-on-entry' feature. Users may
now ask the debugger to halt any new wavefront executing on an AMD GPU using:
(gdb) set amdgpu trap-on-entry on
As soon as the inferior runs a compute workload, control is returned to the
debugger:
Thread 15 "kernel" received signal SIGTRAP, Trace/breakpoint trap.
[Switching to thread 15, lane 0 (AMDGPU Lane 1:2:2:9/0 (0,0,0)[0,0,0])]
kernel () at /home/sdarche/test-dbg/src/app/long_wave.cpp:12
12 __global__ void kernel() {
(gdb)
This feature can be similarly disabled using `set amdgpu trap-on-entry off`.
It is disabled by default.
New testcases are provided in gdb.rocm/trap-on-entry.exp. These testcases
exercise the feature and verify that it disables cleanly.
sebdar
force-pushed
the
users/sdarche/trap-on-entry-support
branch
from
August 10, 2026 20:37
eb00ed3 to
82392de
Compare
Collaborator
|
New PR is ROCm/rocm-systems#9925. Pre-commit seems unhappy about whitespaces. |
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
4 participants
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Motivation
This commit introduces interface to the new 'trap-on-entry' feature. Users may now ask the debugger to halt any new wavefront executing on an AMD GPU using:
(gdb) set amdgpu trap-on-entry on
As soon as the inferior runs a compute workload, control is returned to the debugger:
This feature can be similarly disabled using
set amdgpu trap-on-entry off. It is disabled by default.Tracking
JIRA ID : AIROCGDB-647
Test Plan
New testcases are provided in gdb.rocm/trap-on-entry.exp. These testcases exercise the feature and verify that it disables cleanly.
Test Result
Tests should pass with the new dbgapi. They are unsupported for version < 0.81.
Submission Checklist