[Date Prev][Date Next] [Thread Prev][Thread Next] [Date Index] [Thread Index]

Re: Bug#1081225: gdb: enable AMD GPU support



Hi Emanuele,

On 2024-09-10 09:20, Emanuele Rocca wrote:
Is it sufficient to just add amd-dbgapi-dev to build-depends in
order to get a gdb with AMD GPU debugging support? I tried with
with `sbuild --dist sid gdb --add-depends=amd-dbgapi-dev`, the build
went fine, but then I can't use the commands mentioned on [1], eg.
`show amdgpu precise-memory` does not seem to work.

Is there anything else to be done in order to test the functionality?

Apologies. I'd tested the gdb integration a while ago and I think I forgot I needed additional build options. I have now built and tested it again using the upstream release. Unfortunately, some of the options I've needed are not going to be acceptable for the Debian package. This is how I built and tested it (on a workstation with a gfx906 AMD GPU):

apt install git hipcc amd-dbgapi-dev
git clone https://sourceware.org/git/binutils-gdb.git
cd binutils-gdb
git checkout gdb-15.1-release
apt build-dep gdb
./configure --with-amd-dbgapi=yes --disable-ld --disable-gas --disable-gdbtk --disable-gprofng --disable-shared --disable-sim --enable-targets="x86_64-linux-gnu,amdgcn-amd-amdhsa"
make -j16
make install
cd ..
cat > main.hip << END
#include <stdio.h>
#include <stdlib.h>
#include <hip/hip_runtime.h>

#define CHECK_HIP(expr) do {              \
  hipError_t result = (expr);             \
  if (result != hipSuccess) {             \
    fprintf(stderr, "%s:%d: %s (%d)\n",   \
      __FILE__, __LINE__,                 \
      hipGetErrorString(result), result); \
    exit(EXIT_FAILURE);                   \
  }                                       \
} while(0)

__global__ void sq_arr(float *arr, int n) {
  int tid = blockDim.x*blockIdx.x + threadIdx.x;
  if (tid < n) {
    arr[tid] = arr[tid] * arr[tid];
  }
}

int main() {
  enum { N = 5 };
  float hArr[N] = { 1, 2, 3, 4, 5 };
  float *dArr;
  CHECK_HIP(hipMalloc(&dArr, sizeof(float) * N));
  CHECK_HIP(hipMemcpy(dArr, hArr, sizeof(float) * N, hipMemcpyHostToDevice));
  sq_arr<<<dim3(1), dim3(32,1,1), 0, 0>>>(dArr, N);
  CHECK_HIP(hipMemcpy(hArr, dArr, sizeof(float) * N, hipMemcpyDeviceToHost));
  for (int i = 0; i < N; ++i) {
    printf("%f\n", hArr[i]);
  }
  CHECK_HIP(hipFree(dArr));
  return 0;
}
END
cat > sample.gdb << END
start
next
next
next
print dArr[3]
next
next
print dArr[3]
continue
quit
END
hipcc -g -O0 main.hip
gdb -x sample.gdb ./a.out


The output should look like:

GNU gdb (GDB) 15.1
Copyright (C) 2024 Free Software Foundation, Inc.
License GPLv3+: GNU GPL version 3 or later <http://gnu.org/licenses/gpl.html>
This is free software: you are free to change and redistribute it.
There is NO WARRANTY, to the extent permitted by law.
Type "show copying" and "show warranty" for details.
This GDB was configured as "x86_64-pc-linux-gnu".
Type "show configuration" for configuration details.
For bug reporting instructions, please see:
<https://www.gnu.org/software/gdb/bugs/>.
Find the GDB manual and other documentation resources online at:
    <http://www.gnu.org/software/gdb/documentation/>.

For help, type "help".
Type "apropos word" to search for commands related to "word"...
Reading symbols from ./a.out...
Temporary breakpoint 1 at 0x12d2: file main.hip, line 16.
[Thread debugging using libthread_db enabled]
Using host libthread_db library "/lib/x86_64-linux-gnu/libthread_db.so.1".

Temporary breakpoint 1, main () at main.hip:16
warning: Source file is more recent than executable.
16      float hArr[N] = { 1, 2, 3, 4, 5 };
18      CHECK_HIP(hipMalloc(&dArr, sizeof(float) * N));
[New Thread 0x7fffe6c006c0 (LWP 908081)]
[New Thread 0x7ffee60006c0 (LWP 908082)]
[Thread 0x7ffee60006c0 (LWP 908082) exited]
19      CHECK_HIP(hipMemcpy(dArr, hArr, sizeof(float) * N, hipMemcpyHostToDevice));
20      sq_arr<<<dim3(1), dim3(32,1,1), 0, 0>>>(dArr, N);
$1 = 4
[New Thread 0x7ffed7e006c0 (LWP 908083)]
[Thread 0x7ffed7e006c0 (LWP 908083) exited]
[New Thread 0x7ffff23af6c0 (LWP 908084)]
21      CHECK_HIP(hipMemcpy(hArr, dArr, sizeof(float) * N, hipMemcpyDeviceToHost));
22      for (int i = 0; i < N; ++i) {
$2 = 16
1.000000
4.000000
9.000000
16.000000
25.000000
[Thread 0x7ffff23af6c0 (LWP 908084) exited]
[Thread 0x7fffe6c006c0 (LWP 908081) exited]
[Inferior 1 (process 908078) exited normally]

In summary, it appears there is more work to be done in fixing the incompatibilities with other gdb features before this option can be enabled. I was also browsing the upstream repository and I noticed changes that require the most recent release of rocdbgapi [1]. We'll probably have ROCm 6.2 packaged by the time gdb 16 is released, but we'll have to keep that in mind.

Sincerely,
Cory Bloor

[1]: https://sourceware.org/git/?p=binutils-gdb.git;a=commit;h=37ef6d976a777f186568f2b629f04a6bd9706ed7


Reply to: