[PATCH 4/6] gdb/linux-tdep: don't write core file notes for non-Linux threads

[Simon Marchi <simon.marchi@efficios.com>]

Running "gcore" while debugging a ROCm program, with some GPU waves
stopped, makes GDB crash:

  (gdb) gcore
  .../gdb/gdbarch-gen.c:3637: internal-error: gdbarch_iterate_over_regset_sections: Assertion `gdbarch->iterate_over_regset_sections != nullptr' failed.

linux_make_corefile_notes writes register notes for every thread of the
inferior, using each thread's own gdbarch.  When debugging a ROCm
program, the inferior's thread list contains both the host (Linux)
threads and the GPU waves, provided by the amd-dbgapi target.  The waves
use the amdgcn gdbarch, which doesn't implement
iterate_over_regset_sections, hence the assertion failure.  This happens
regardless of which thread is selected, since all threads are visited.

There is also a related problem: if a wave is the signalled thread, for
instance because it is the selected thread and hit a breakpoint, it is
dumped first, its stop signal is used for all the other threads, and its
gdbarch is used to write the NT_GDB_TDESC note.  This would produce a
core file with an amdgcn target description.

Make linux_make_corefile_notes skip non-Linux threads, and only consider
Linux threads when looking for the signalled thread.

To identify Linux threads, add the is_linux_thread function.  It relies
on the fact that the amd-dbgapi target uses ptids of the form (pid, 1,
wave_id), with a non-zero tid field, while Linux threads have a zero tid
field.  This is not super pretty, but we need to teach the Linux support
that there might be non-Linux threads in the thread list one way or
another, and this is cheap, simple and reliable.  If we ever find a
better way to discriminate threads, then we just have to update this
function.

Add an optional filter parameter to gcore_find_signalled_thread, to
let linux_make_corefile_notes exclude non-Linux threads.

Add the gdb.rocm/gcore-host.exp test, which generates a core file while
a GPU wave is stopped at a breakpoint, and a host thread is selected.
It then loads the core file and checks that it contains all the host
threads, and only them.  Generating a core file while a GPU wave is
selected still fails, because of a separate bug that is addressed in a
following patch.

The test is named like that to avoid clashing with the
gdb.rocm/gcore.exp test downstream, which tests core generation
including the GPU threads.  This feature is going to be upstreamed
eventually, but in the meantime the gdb.rocm/gcore-host.exp test is
meant to test that the presence of GPU waves does not prevent a
host-only core dump from being created.

Change-Id: I33fd7fcf21ca8c645195a821e8dc65f042cda364
---
 gdb/gcore.c                           |  27 ++++---
 gdb/gcore.h                           |   9 ++-
 gdb/linux-tdep.c                      |  29 +++++++-
 gdb/testsuite/gdb.rocm/gcore-host.cpp |  42 +++++++++++
 gdb/testsuite/gdb.rocm/gcore-host.exp | 102 ++++++++++++++++++++++++++
 5 files changed, 195 insertions(+), 14 deletions(-)
 create mode 100644 gdb/testsuite/gdb.rocm/gcore-host.cpp
 create mode 100644 gdb/testsuite/gdb.rocm/gcore-host.exp

diff --git a/gdb/gcore.c b/gdb/gcore.c
index 77d915505d1c..1336e56cda82 100644
--- a/gdb/gcore.c
+++ b/gdb/gcore.c
@@ -903,22 +903,31 @@ gcore_memory_sections (bfd *obfd)
 /* See gcore.h.  */
 
 thread_info *
-gcore_find_signalled_thread ()
+gcore_find_signalled_thread
+  (gdb::function_view<bool (const thread_info &)> filter)
 {
+  auto accept = [filter] (const thread_info &thr)
+    {
+      return filter == nullptr || filter (thr);
+    };
+
   thread_info *curr_thr = inferior_thread ();
-  if (curr_thr->state () != THREAD_EXITED
-      && curr_thr->stop_signal () != GDB_SIGNAL_0)
+
+  /* Give priority to the current thread, unless it has exited or is not
+     accepted by FILTER.  */
+  if (curr_thr->state () == THREAD_EXITED || !accept (*curr_thr))
+    curr_thr = nullptr;
+
+  if (curr_thr != nullptr && curr_thr->stop_signal () != GDB_SIGNAL_0)
     return curr_thr;
 
   for (thread_info &thr : current_inferior ()->non_exited_threads ())
-    if (thr.stop_signal () != GDB_SIGNAL_0)
+    if (accept (thr) && thr.stop_signal () != GDB_SIGNAL_0)
       return &thr;
 
-  /* Default to the current thread, unless it has exited.  */
-  if (curr_thr->state () != THREAD_EXITED)
-    return curr_thr;
-
-  return nullptr;
+  /* Default to the current thread, unless it has exited or is not
+     accepted by FILTER, in which case return nullptr.  */
+  return curr_thr;
 }
 
 INIT_GDB_FILE (gcore)
diff --git a/gdb/gcore.h b/gdb/gcore.h
index 2431e123b60e..277c28d9c9ac 100644
--- a/gdb/gcore.h
+++ b/gdb/gcore.h
@@ -22,6 +22,7 @@
 
 #include "find-memory-region.h"
 #include "gdb_bfd.h"
+#include "gdbsupport/function-view.h"
 
 struct thread_info;
 
@@ -33,8 +34,12 @@ extern bool objfile_find_memory_regions (struct target_ops *self,
 /* Find the signalled thread.  In case there's more than one signalled
    thread, prefer the current thread, if it is signalled.  If no thread was
    signalled, default to the current thread, unless it has exited, in which
-   case return NULL.  */
+   case return nullptr.
 
-extern thread_info *gcore_find_signalled_thread ();
+   If FILTER is non-null, only consider threads for which FILTER returns
+   true.  */
+
+extern thread_info *gcore_find_signalled_thread
+  (gdb::function_view<bool (const thread_info &)> filter = nullptr);
 
 #endif /* GDB_GCORE_H */
diff --git a/gdb/linux-tdep.c b/gdb/linux-tdep.c
index fbc96c0975a7..3491c63c017a 100644
--- a/gdb/linux-tdep.c
+++ b/gdb/linux-tdep.c
@@ -456,6 +456,26 @@ linux_has_shared_address_space (struct gdbarch *gdbarch)
   return linux_is_uclinux ();
 }
 
+/* Return true if THREAD appears to be a Linux thread.
+
+   When debugging a ROCm program, an inferior contains both host threads
+   (provided by the linux-nat target) and GPU threads (provided by the
+   amd-dbgapi target).
+
+   To differentiate them from the host threads, the amd-dbgapi target uses
+   ptids of the form
+
+     (pid, 1, wave_id)
+
+   where wave_id is a non-zero unique number.  Return true (is a Linux
+   thread) if the tid field is 0.  */
+
+static bool
+is_linux_thread (const thread_info &thread)
+{
+  return thread.ptid.tid () == 0;
+}
+
 /* Return the ptid of a thread of the current inferior whose lwp id can be
    used to build "/proc/<lwp>" paths.
 
@@ -2524,8 +2544,11 @@ linux_make_corefile_notes (struct gdbarch *gdbarch, bfd *obfd, int *note_size)
 
   /* Like the kernel, prefer dumping the signalled thread first.
      "First thread" is what tools use to infer the signalled
-     thread.  */
-  thread_info *signalled_thr = gcore_find_signalled_thread ();
+     thread.
+
+     Only consider Linux threads.  Other threads, like AMD GPU waves, can't be
+     described using Linux thread notes.  */
+  thread_info *signalled_thr = gcore_find_signalled_thread (is_linux_thread);
   gdb_signal stop_signal;
   if (signalled_thr != nullptr)
     stop_signal = signalled_thr->stop_signal ();
@@ -2546,7 +2569,7 @@ linux_make_corefile_notes (struct gdbarch *gdbarch, bfd *obfd, int *note_size)
     }
   for (thread_info &thr : current_inferior ()->non_exited_threads ())
     {
-      if (&thr == signalled_thr)
+      if (&thr == signalled_thr || !is_linux_thread (thr))
 	continue;
 
       /* On some architectures, like AArch64, each thread can have a distinct
diff --git a/gdb/testsuite/gdb.rocm/gcore-host.cpp b/gdb/testsuite/gdb.rocm/gcore-host.cpp
new file mode 100644
index 000000000000..0f8ca14fe90f
--- /dev/null
+++ b/gdb/testsuite/gdb.rocm/gcore-host.cpp
@@ -0,0 +1,42 @@
+/* This testcase is part of GDB, the GNU debugger.
+
+   Copyright 2026 Free Software Foundation, Inc.
+
+   This program is free software; you can redistribute it and/or modify
+   it under the terms of the GNU General Public License as published by
+   the Free Software Foundation; either version 3 of the License, or
+   (at your option) any later version.
+
+   This program is distributed in the hope that it will be useful,
+   but WITHOUT ANY WARRANTY; without even the implied warranty of
+   MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
+   GNU General Public License for more details.
+
+   You should have received a copy of the GNU General Public License
+   along with this program.  If not, see <http://www.gnu.org/licenses/>.  */
+
+#include "hip/hip_runtime.h"
+#include <cassert>
+
+__global__ void
+kernel (int *out)
+{
+  *out = 1;
+}
+
+int
+main ()
+{
+  int *result_ptr, result;
+
+  hipError_t error = hipMalloc (&result_ptr, sizeof (int));
+  assert (error == hipSuccess);
+
+  kernel<<<dim3 (1), dim3 (1), 0, 0>>> (result_ptr);
+
+  error = hipMemcpyDtoH (&result, result_ptr, sizeof (int));
+  assert (error == hipSuccess);
+  assert (result == 1);
+
+  return 0;
+}
diff --git a/gdb/testsuite/gdb.rocm/gcore-host.exp b/gdb/testsuite/gdb.rocm/gcore-host.exp
new file mode 100644
index 000000000000..80bf60e4e9f5
--- /dev/null
+++ b/gdb/testsuite/gdb.rocm/gcore-host.exp
@@ -0,0 +1,102 @@
+# Copyright 2026 Free Software Foundation, Inc.
+
+# This program is free software; you can redistribute it and/or modify
+# it under the terms of the GNU General Public License as published by
+# the Free Software Foundation; either version 3 of the License, or
+# (at your option) any later version.
+#
+# This program is distributed in the hope that it will be useful,
+# but WITHOUT ANY WARRANTY; without even the implied warranty of
+# MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
+# GNU General Public License for more details.
+#
+# You should have received a copy of the GNU General Public License
+# along with this program.  If not, see <http://www.gnu.org/licenses/>.
+
+# Test generating a core file for an inferior that has GPU waves, then
+# loading it.  GDB does not support writing GPU waves to core files, so the
+# core file is expected to contain only the host threads.
+
+load_lib rocm.exp
+
+standard_testfile .cpp
+
+require allow_hip_tests
+
+if {[build_executable "failed to prepare" $testfile $srcfile {debug hip}]} {
+    return
+}
+
+# Generate a core file while a GPU wave is stopped at a breakpoint, with
+# SELECTED_THREAD ("host" or "wave") being the selected thread.  Then load
+# the core file and verify its contents.
+
+proc do_test { selected_thread } {
+    clean_restart $::testfile
+
+    set corefile [standard_output_file $::testfile.$selected_thread.core]
+
+    with_rocm_gpu_lock {
+	if {![runto_main]} {
+	    return
+	}
+
+	gdb_breakpoint "kernel" -allow-pending
+
+	gdb_test "continue" \
+	    "Thread $::decimal hit Breakpoint ${::decimal}(?:\\.${::decimal})?, kernel .*"
+
+	# Count the host threads, for comparison with the core file.
+	set num_host_threads 0
+	set num_waves 0
+	gdb_test_multiple "info threads" "" -lbl {
+	    -re "\r\n\\*? +$::decimal +Thread $::hex \\(LWP $::decimal\\)\[^\r\n\]*" {
+		incr num_host_threads
+		exp_continue
+	    }
+	    -re "\r\n\\*? +$::decimal +AMDGPU Wave\[^\r\n\]*" {
+		incr num_waves
+		exp_continue
+	    }
+	    -re -wrap "" {
+		gdb_assert { $num_host_threads > 0 && $num_waves > 0 } \
+		    $gdb_test_name
+	    }
+	}
+
+	if { $selected_thread == "host" } {
+	    gdb_test "thread 1" "Switching to thread 1 .*"
+	}
+
+	if { [gdb_gcore_cmd $corefile "save a corefile"] != 1 } {
+	    return
+	}
+    }
+
+    clean_restart $::testfile
+
+    gdb_test "core-file $corefile" \
+	"Core was generated by `[string_to_regexp $::binfile]'\\..*" \
+	"load corefile"
+
+    # The core file is expected to contain all the host threads, and nothing
+    # else.
+    set num_core_threads 0
+    gdb_test_multiple "info threads" "core file contains host threads" -lbl {
+	-re "\r\n\\*? +$::decimal +Thread $::hex \\(LWP $::decimal\\)\[^\r\n\]*" {
+	    incr num_core_threads
+	    exp_continue
+	}
+	-re "\r\n\\*? +$::decimal +\[^\r\n\]*" {
+	    fail $gdb_test_name
+	}
+	-re -wrap "" {
+	    gdb_assert { $num_core_threads == $num_host_threads } \
+		$gdb_test_name
+	}
+    }
+}
+
+foreach_with_prefix selected_thread { host } {
+    do_test $selected_thread
+}
-- 
2.56.0