gdb/testsuite: support optimized code in gdb.rocm testsuite

[spatrang]

Summary

This series makes the gdb.rocm testsuite pass when the test programs are
compiled with optimization (-O1/-O2/-O3, -flto) passed via
GDB_TESTCASE_OPTIONS, rather than the implicit -O0 the suite assumes today.

Three commits, split by concern:

1. Update gdb.rocm framework to support optimized code — adds helpers in
   gdb/testsuite/lib/rocm.exp:

   • check_for_any_flags {flags} — detects whether GDB_TESTCASE_OPTIONS
     contains any of the listed flags.
   • gdb_caching_proc no_optimized_code — predicate for -O1/-O2/-O3.
   • gdb_caching_proc no_lto — predicate for -flto.

2. Update some gdb.rocm tests to support optimized code (22 files) —
   two patterns applied surgically:

   • In .cpp/.c: add volatile on locals the test inspects, and
     __attribute__((optnone)) on helper functions the test breakpoints into.
     Add a trivial volatile int statement to empty "break here" functions
     so their body is not elided.
   • In .exp: relax line-number arithmetic, backtrace PC formatting,
     expected source strings after step, and add a bounded retry loop for
     extra ticks under optimization.

3. Skip unsupported tests in case of optimized code (19 files) — for tests
   whose semantics cannot be reconciled with optimized code, add
   require no_optimized_code at the top; mi-aspace.exp also gets
   require no_lto. register-watchpoint.exp soft-skips the tail via
   check_for_any_flags after the earlier portion has run.

Diffstat: 42 files changed, +146 / -46.

Context

This continues the work started in internal PR
AMD-ROCm-Internal/ROCgdb#6,
rebased onto the current amd-staging of the public repo. The commits are
content-identical to the internal PR, preserving the original authors
(Aleksandar Rikalo, Andrey Kasaurov) with their Co-Authored-By: trailers
and adding my Signed-off-by.

[spatrang]

Hi @lancesix

This PR is a re-submission of the optimized-code series that was originally
opened in the internal rocgdb repository. I'm summarising your comments and
how they've been addressed in the latest push:

1. gdb.rocm/deep-stack.exp — "This part of the change seems odd. Is
   it really necessary?"
   Reverted. Frames #10..#12 now use the same ".* <name>" regex
   style as #0..#9; the dropped space was unnecessary.

2. gdb.rocm/device-attach.cpp — "Is code still valid without this
   line?"
   No, it isn't. Restored the __global__ void qualifier on
   bit_extract_kernel; without it the function isn't a HIP kernel
   and hipLaunchKernelGGL would fail to compile.

3. gdb.rocm/hip-builtin-variables.cpp — "Not a fan of the implicit
   int … const volatile reads off … Looks like it could just be
   volatile int."
   Agreed. All const volatile X = …; declarators have been replaced
   with volatile int X = …;.

4. gdb.rocm/precise-memory-exec.c — "Why volatile int sometimes,
   attribute((optnone)) other times?"
   Good point. second() now uses __attribute__((optnone)) instead
   of an artificial volatile int i = 0;, matching the convention
   used in corefile.cpp, names.cpp, ocp_mx.cpp, and
   step-schedlock-spurious-waves.cpp (touched in the same series).

5. lib/rocm.exp — "missing '.'" (×2) and "indentation" (×2)
   on no_optimized_code and no_lto.
   Both comments were added a trailing period, and the body of the
   if-blocks in both procs is now tab-indented to match the
   surrounding style.

The corresponding fix-up commit is on top of this branch. Please take
another look when you have a moment

cc: @lumachad

[lumachad]

pre-commit seems to be grumpy.

[lumachad]

As stated elsewhere, the series only handles GDB_TESTCASE_OPTIONS. If someone passes CFLAGS directly to the compiler the optimization level checks won't work properly. Do we want to fix that as well? Maybe as a follow up if not in this one?

[lumachad]

There's a pre-commit failure.

Other than that, I'm still seeing FAIL's for gdb.rocm/watchpoint-basic.exp when running the testsuite with -O1 or -O3. It passes when running with -O0. Could you please check that you're running the testsuite correctly?

[lumachad]

Did the latest force-pushes provide anything that would fix gdb.rocm/watchpoint-basic.exp? I can't tell.

[lumachad]

Just small issues.

[lumachad]

For -lto runs, it looks like gdb.rocm/hip-builtin-variables.exp and gdb.rocm/hip-lang-detect.exp are still failing.

FAIL: gdb.rocm/hip-lang-detect.exp: language detection (dwarfdump execution)
FAIL: gdb.rocm/hip-lang-detect.exp: expected compiler language (found "unknown")

[spatrang]

For -lto runs, it looks like gdb.rocm/hip-builtin-variables.exp and gdb.rocm/hip-lang-detect.exp are still failing.

FAIL: gdb.rocm/hip-lang-detect.exp: language detection (dwarfdump execution)
FAIL: gdb.rocm/hip-lang-detect.exp: expected compiler language (found "unknown")

Both hip-builtin-variables.exp and hip-lang-detect.exp now require no_lto. LTO can re-inline across TUs (defeating noinline/optnone boundaries) and rewrite the per-CU DW_LANG attribute, so the test premises don't hold at -flto. Matches the existing require no_lto pattern in watchpoint-at-end-of-shader.exp and mi-aspace.exp.

[lumachad]

Only had a minor comment. Otherwise this LGTM. I'd like @lancesix to eyeball this as well.

[lumachad]

Things look clean from -O1 through -O3. I do see a FAIL with -flto though:

FAIL: gdb.rocm/hip-builtin-completions.exp: language detection (dwarfdump execution)

Since we claim to be clean on -flto runs as well, we should fix that or skip it.

[lumachad]

Confirmed a clean run for -flto as well. Thanks.

[lumachad]

Leaving my approval here. If Lancelot doesn't have any restrictions on this series, feel free to merge.

[lancesix]

Comments on patch 1.

[lancesix]

Comments for patch 2.

It seems that this pattern is not consistently implemented

• Make breakpoint-anchor helper functions ("done", "foo",
  "bar") non-static and noinline, with an asm volatile body,
  so the symbol survives the optimiser and the breakpoint
  can be inserted.

Seems that some of that logic could be factored out to help (would avoid having to repeat the comment giving the rationale for the asm and so on).

[lancesix]

Comments for the 3rd path.

It really seems that many of the changes from this patch belong to the previous one (many changes are just modifying what the previous patch). Arguably the previous patch does too much at once and could be split, but regardless the fixup should not end-up here.

Seems like Lancelot has more comments. I'll let him take over.

[spatrang]

Hello @lancesix @lumachad
Pushed an updated revision addressing the latest review round:

• Split out the deep-stack.exp backtrace regex into its own commit (it's an independent fix from the optimization work).
• Folded the "belongs in previous patch" churn back into the right commits (no more add-then-revert across commits, e.g. deref-scoped-pointer, corefile).
• rocm.exp: check_for_any_flags/no_optimized_code/no_lto now use conventional boolean returns (1/0) and inline if {[check_for_any_flags ...]}; dropped the happy-case log lines.
• Style: attribute ((...)) spacing in corefile.cpp and finish.cpp.
• alu-exceptions / fork-exec-* / precise-memory-*: dropped the artificial volatile int locals in favor of noinline + asm volatile ("" ::: "memory").
• lane-info.cpp: fixed the UB infinite loop (volatile bool keep_going); kept noinline where the test needs a real frame.
• step-schedlock-spurious-waves.exp: merged the redundant retry loop into the existing max_iter budget.
• hip-builtin-completions.exp: added require no_lto (the dwarfdump language check is unreliable under -flto).

Re-tested the modified tests at -O0/-O1/-O2/-O3: 0 failures (the 3 UNSUPPORTED at -O1+ are the intended watchpoint-basic before-runtime-load skips).