gdb: add "catch hiperr" to break on HIP API errors

[amd-shahab]

This change relies on "Add __hipOnError debugger hook for HIP API failures" PR to be functional. Although, without it, ROCgdb won't be broken. Please read the commit message for the description of this change itself.

--

What

This change adds a new feature to GDB, when it is targeted for
amdgcn-amd-amdhsa, that enables it to stop whenever HIP API
errors are returned. An abridged sample of execution looks
like this:

(gdb) catch hiperr
(gdb) run
Thread 1 "hip" hit Catchpoint 1 (HIP error)
hipErrorInvalidDevice (101): invalid device ordinal

Moreover, a $_hiperr convenience variable is added that holds
the error code. For further details, please see the updated
documentation in this change.

Design

Most of the implementation is in the newly added module called
break-catch-hiperr.c. The naming and its context follows the
break-catch-throw.c module.

The whole mechanism work by setting a breakpoint (catchpoint)
on a __hipOnError() symbol, that must be provided by the CLR
library. Enabling the catchpoint when there's no __hipOnError
symbol, simply has no effect.

There's a architect dependent section that is responsible for
extracting the error parameters (code, name, description) from
a single argument passed to the __hipOnError() function as a
structure pointer. The expected format of this struct is:

uint32_t version; // at this point, only 1 is supported
uint32_t code; // the numerical error code
const char *name; // name of the error
const char *desc; // a phrase about the error

The code is devised in such a way that if the __hipOnError()
exists, but for whatever reason the parameters cannot be deduced,
it won't interfere with the catchpoint being triggered. It will
affect the reported data of the catchpoint though.

Test

A new test, gdb.rocm/hip-catch-errors, is added to the mix to
ensure that:

1. (Pending) catchpoints work.
2. Multiple catchpoints in one execution are reported accordingly.
3. Errors of indirect HIP calls like "kenel<<<...>>>()" are caught.
4. The convenience variable can be used for conditional situations.
5. Conditional and temporary catchpoints work as expected.

ROCM-1548

[lumachad]

@amd-shahab If this shouldn't be tested yet and isn't mergeable until the dependent PR, feel free to label it "ci:skip".

[lancesix]

First quick first pass of comments.

One question regarding the test: what kind of guarantee do we have that the error code and error string are going to be stable over time? Let's imagine that the hip runtime decides to localize the error string (not the name). Couldn't that be a legitimate thing to do, but it would break the test, would it?

[amd-shahab]

• error string not part of API

• __hipOnError -> @samp{__hipOnError}

• "int main" -> "int\nmain"

• gdb_test -> gdb_test_on_output

• target_read_stack -> target_read_memory

• drop the "ignored" param name

• Else -> Otherwise

• NULL -> nullptr

• removed dangling "backtrace" command from the documentation

• !arg -> arg != nullptr

• binay -> binary

• relvant -> relevant

• bool print_one (const bp_location **) -> bool print_one (const bp_location **last_loc)

• rename variable "x" to "dontcare"

• comment on true/false return of amd64_fetch_hiperr_parameters()

[amd-shahab]

• emit a warning when the "version" is not expected.

[amd-shahab]

• make the invocation of the test fully argument dependent
• add comments for every method in break-catch-hiperr.c

[amd-shahab]

• check for the existence of __hipOnError symbol. If not found, skip the test.

[amd-shahab]

• check for the right frame (__hipOnError)

[amd-shahab]

• $_hiperr prints as hipError_t if that type is known to GDB.
• Add tests for that.
• Update the documentation to reflect that.

[amd-shahab]

• Add test for when there's no debug symbols

[amd-shahab]

• remove static from static struct block_symbol bs definition, so the symbol will always be looked up.

[palves]

In commit log, drop the "when it is targeted for amdgcn-amd-amdhsa" clause from the first sentence.

[amd-shahab]

• remove "amdgcn target" in commit log
• put $_hiperr example with conditional breakpoint
• test: do not match thread name so things work on Windows
• doc: reword the first paragraph
• comments: remove "supposedly"
• fix a few typos

[palves]

• test: do not match thread name so things work on Windows

AFAICS, you're still matching the thread name. The thread name is the thing in double quotes.

[amd-shahab]

• check hipError_t is a 4-byte enum.

[amd-shahab]

• test: do not match thread name so things work on Windows

AFAICS, you're still matching the thread name. The thread name is the thing in double quotes.

Now the pattern is:

set hit_line "Thread $::decimal .* hit $type $::decimal \\(HIP error\\)"

[amd-shahab]

• test: replace the double-quoted part in hit_line with an any (.*).
• test: fix indentation of a closing } for the foreach

[amd-shahab]

• doc: add hiperr to catch section of the documentation. cross-linked by the "HIP API Errors" subsection.

[amd-shahab]

• test: spell out the whole header line for the output of info types hipError_t.

[amd-shahab]

• doc: drop [if condition...] part from catch hiperr entry.
• doc: reword the catch hiperr section to the suggested paragraph.
• doc: add @cindex for HIP runtime error to go with HIP runtime error convenience variable.