Skip to content
This repository has been archived by the owner on Mar 21, 2024. It is now read-only.

Add GDB pretty-printer for thrust vectors and references #1631

Merged
merged 2 commits into from
May 5, 2022

Conversation

upsj
Copy link
Contributor

@upsj upsj commented Mar 16, 2022

I saw #1318 and really liked the idea of providing pretty-printers even for device-side data, so I implemented it in our own library (ginkgo-project/ginkgo#987), and thought it might be useful here as well :) I removed all non-trivial parts of the code that were copied from libstdc++, so there should be no licensing issues. This is should be straightforward to extend to device references , I only had a few issues with getting the device pointer out from a device_reference (gdb returns a heap pointer for D[0].ptr.m_iterator, but what looks like a device pointer for D.m_storage.m_begin.m_iterator.m_iterator in examples/basic_vector.cu, so there must be something weird going wrong with the temporary D[0] I guess?)

I can't really speak on the potential for hangs that was discussed in the original issue, aside from "it works for me" - that would be something the runtime folks need to answer.

Example run:

(gdb) b 33
Breakpoint 1 at 0xc3e8: file /home/tobi/thrust/examples/basic_vector.cu, line 33.
(gdb) run
Starting program: /home/tobi/thrust/build/bin/thrust.example.basic_vector 
[Thread debugging using libthread_db enabled]
Using host libthread_db library "/lib/x86_64-linux-gnu/libthread_db.so.1".
H has size 4
H[0] = 14
H[1] = 20
H[2] = 38
H[3] = 46
H now has size 2
[New Thread 0x7fffefea3000 (LWP 2198)]

Thread 1 "thrust.example." hit Breakpoint 1, main () at /home/tobi/thrust/examples/basic_vector.cu:33
33          D[0] = 99;
(gdb) source ../scripts/gdb-pretty-printers.py 
(gdb) print H
$1 = thrust::host_vector<int, std::allocator<int> > of length 2, capacity 4 = {14, 20}
(gdb) print D
$2 = thrust::device_vector<int, thrust::device_allocator<int> > of length 2, capacity 2 = {14, 20}
(gdb) print D[1]
$3 = (thrust::device_reference<int>) @0xb047f0004: 20
(gdb)

Fixes #1318

@alliepiper alliepiper added type: enhancement New feature or request. P2: nice to have Desired, but not necessary. labels Mar 16, 2022
@alliepiper alliepiper added this to the 1.17.0 milestone Mar 16, 2022
@alliepiper
Copy link
Collaborator

Awesome, thanks! We'll take a look soon, but it may be sometime after GTC.

@GPUtester
Copy link
Collaborator

Can one of the admins verify this patch?

@upsj upsj changed the title Add GDB pretty-printer for thrust vectors Add GDB pretty-printer for thrust vectors and references Apr 30, 2022
@upsj
Copy link
Contributor Author

upsj commented Apr 30, 2022

I figured out the issues with device references, turns out GDB <= 9.2 has some bug I didn't bother to track down that has been fixed in GDB 10.1, so we can actually conditionally add pretty printers for references as well :)

Copy link
Collaborator

@ericniebler ericniebler left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Looks good to me. Thanks!

@ericniebler ericniebler merged commit 48d995c into NVIDIA:main May 5, 2022
@upsj upsj deleted the pretty_printer branch May 5, 2022 18:19
rapids-bot bot pushed a commit to rapidsai/rmm that referenced this pull request Aug 26, 2022
This PR adds a pretty-printer for `device_uvector` and pulls in Thrust pretty-printers that were added in NVIDIA/thrust#1631. CMake provides a convenience script to load all of the pretty-printers, to resolve the duplication concerns raised in rapidsai/cudf#11499.

Example output:
<details>

```
$ gdb -q gtests/DEVICE_UVECTOR_TEST
Reading symbols from gtests/DEVICE_UVECTOR_TEST...
(gdb) b cudaMalloc
Function "cudaMalloc" not defined.
Make breakpoint pending on future shared library load? (y or [n]) y
Breakpoint 1 (cudaMalloc) pending.
(gdb) run
Starting program: /home/nfs/tribizel/rapids/rmm/build/cuda-11.5.0/feature__pretty-printers/debug/gtests/DEVICE_UVECTOR_TEST 
warning: Error disabling address space randomization: Operation not permitted
[Thread debugging using libthread_db enabled]
Using host libthread_db library "/usr/lib/x86_64-linux-gnu/libthread_db.so.1".
Running main() from gmock_main.cc
[==========] Running 95 tests from 5 test suites.
[----------] Global test environment set-up.
[----------] 19 tests from TypedUVectorTest/0, where TypeParam = signed char
[ RUN      ] TypedUVectorTest/0.MemoryResource
[New Thread 0x7f741efc1000 (LWP 86147)]

Thread 1 "DEVICE_UVECTOR_" hit Breakpoint 1, 0x00007f7426dddc80 in cudaMalloc () from /home/nfs/tribizel/rapids/compose/etc/conda/cuda_11.5/envs/rapids/lib/libcudart.so.11.0
(gdb) c
Continuing.
[New Thread 0x7f741e7c0000 (LWP 86148)]
[       OK ] TypedUVectorTest/0.MemoryResource (999 ms)
[ RUN      ] TypedUVectorTest/0.ZeroSizeConstructor
[       OK ] TypedUVectorTest/0.ZeroSizeConstructor (0 ms)
[ RUN      ] TypedUVectorTest/0.NonZeroSizeConstructor

Thread 1 "DEVICE_UVECTOR_" hit Breakpoint 1, 0x00007f7426dddc80 in cudaMalloc () from /home/nfs/tribizel/rapids/compose/etc/conda/cuda_11.5/envs/rapids/lib/libcudart.so.11.0
(gdb) finish
Run till exit from #0  0x00007f7426dddc80 in cudaMalloc () from /home/nfs/tribizel/rapids/compose/etc/conda/cuda_11.5/envs/rapids/lib/libcudart.so.11.0
rmm::mr::cuda_memory_resource::do_allocate (bytes=<optimized out>, this=<optimized out>) at /home/nfs/tribizel/rapids/rmm/include/rmm/mr/device/cuda_memory_resource.hpp:70
70          RMM_CUDA_TRY_ALLOC(cudaMalloc(&ptr, bytes));
(gdb) finish
Run till exit from #0  rmm::mr::cuda_memory_resource::do_allocate (bytes=<optimized out>, this=<optimized out>) at /home/nfs/tribizel/rapids/rmm/include/rmm/mr/device/cuda_memory_resource.hpp:70
0x00005587874feef0 in rmm::device_buffer::allocate_async (bytes=12345, this=0x7ffefebb46b0) at /home/nfs/tribizel/rapids/rmm/include/rmm/device_buffer.hpp:418
418         _data     = (bytes > 0) ? memory_resource()->allocate(bytes, stream()) : nullptr;
Value returned is $1 = (void *) 0x7f73ff000000
(gdb) finish
Run till exit from #0  0x00005587874feef0 in rmm::device_buffer::allocate_async (bytes=12345, this=0x7ffefebb46b0) at /home/nfs/tribizel/rapids/rmm/include/rmm/device_buffer.hpp:418
TypedUVectorTest_NonZeroSizeConstructor_Test<signed char>::TestBody (this=<optimized out>) at /home/nfs/tribizel/rapids/rmm/tests/device_uvector_tests.cpp:55
55        EXPECT_EQ(vec.size(), size);
(gdb) print vec
$2 = {_storage = {_data = 0x7f73ff000000, _size = 12345, _capacity = 12345, _stream = {stream_ = 0x0}, _mr = 0x558787561008 <rmm::mr::detail::initial_resource()::mr>}}
(gdb) source load-pretty-printers 
(gdb) print vec
$3 = rmm::device_uvector<signed char> of length 12345, capacity 12345 = {0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 
  0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000', 0 '\000'...}
(gdb) 
```

</details>

Authors:
  - Tobias Ribizel (https://github.com/upsj)

Approvers:
  - Mark Harris (https://github.com/harrism)
  - Bradley Dice (https://github.com/bdice)

URL: #1088
rapids-bot bot pushed a commit to rapidsai/cudf that referenced this pull request Sep 9, 2022
This adds `gdb` pretty printers for `rmm::device_uvector`, `thrust::*_vector`, `thrust::device_reference` and `cudf::*_span`. The implementation is based on NVIDIA/thrust#1631. I will probably backport the thrust-specific changes to there as well, but since the location of the thrust source is not fixed, I'd prefer having all types in a self-contained file.

Authors:
  - Tobias Ribizel (https://github.com/upsj)

Approvers:
  - Bradley Dice (https://github.com/bdice)
  - Karthikeyan (https://github.com/karthikeyann)

URL: #11499
Sign up for free to subscribe to this conversation on GitHub. Already have an account? Sign in.
Labels
P2: nice to have Desired, but not necessary. type: enhancement New feature or request.
Projects
None yet
Development

Successfully merging this pull request may close these issues.

Provide GDB pretty-printers for Thrust containers and iterators
4 participants