From: Tankut Baris Aktemur <tankut.baris.aktemur@intel.com>
To: gdb-patches@sourceware.org,
Markus Metzger <markus.t.metzger@intel.com>,
config-patches@gnu.org
Subject: [PATCH v3 35/44] gdbserver, ze, intelgt: introduce ze-low and intel-ze-low targets
Date: Fri, 1 Aug 2025 11:37:37 +0200 [thread overview]
Message-ID: <20250801-upstream-intelgt-mvp-v3-35-59ce0f87075b@intel.com> (raw)
In-Reply-To: <20250801-upstream-intelgt-mvp-v3-0-59ce0f87075b@intel.com>
From: Markus Metzger <markus.t.metzger@intel.com>
Add the Level-Zero based intelgt target. It is configured as
intelgt-*-zebin
This adds fundamental debug support for Intel GT devices. In the
future, we plan to add more patches that improve the performance as
well the user experience. Those patches are already available in the
downstream "Intel Distribution for GDB" debugger at
https://github.com/intel/gdb
For Level-Zero based devices, we model hardware threads. There is one
GDB thread for each hardware thread on the device.
We distinguish RUNNING and UNAVAILABLE thread execution states. They
are pretty similar but UNAVAILABLE means that we have tried to stop
the thread and failed, whereas RUNNING means we have resumed the
thread and since not tried to interact with it.
On attach to PID, attach to that PID on all supported devices and
create target descriptions and a process for each device. This
version does not allow attaching to or detaching from and reattaching
to individual devices.
The target can be used in combination with a native target, relying on
GDB's multi-target feature, to debug the GPU and the host application
in the same debug session. For this, bring the native app to a state
where the Level-Zero backend for the GPU has been initialized, then
create a gdbserver instance and connect to it from a second inferior.
Below is a sample session that shows how to do this manually. In the
downstream debugger, a Python script is used to take these steps
in an automated manner for better user experience.
$ gdb demo
...
(gdb) maintenance set target-non-stop on
(gdb) tbreak 60
Temporary breakpoint 1 at 0x4049c8: file demo.cpp, line 60.
(gdb) run
...
[SYCL] Using device: [Intel(R) Arc(TM) A750 Graphics] from [Intel(R) Level-Zero]
Thread 1 "demo" hit Temporary breakpoint 1, main (argc=1, argv=0x7fffffffd9b8) at demo.cpp:60
60 range data_range{length};
(gdb)
# Connect the Intel GT gdbserver by specifying the host inferior PID.
(gdb) add-inferior -no-connection
[New inferior 2]
Added inferior 2
(gdb) inferior 2
[Switching to inferior 2 [<null>] (<noexec>)]
(gdb) info inferiors
Num Description Connection Executable
1 process 16458 1 (native) /temp/demo
* 2 <null>
(gdb) target remote | gdbserver-ze --attach - 16458
Remote debugging using | gdbserver-ze --attach - 16458
Attached; given pid = 16458, updated to 1
Remote debugging using stdio
<unavailable> in ?? ()
(gdb)
# For "continue" to conveniently resume both inferiors,
# set the 'schedule-multi' mode. Then define a breakpoint
# inside the kernel and resume to hit that BP.
(gdb) set schedule-multiple on
(gdb) break demo.cpp:32
Breakpoint 2 at 0x40651f: file demo.cpp, line 32.
(gdb) continue
Continuing.
...
Thread 2.201 hit Breakpoint 2.2, compute (index=sycl::_V1::id<1> = {...},
element=116) at demo.cpp:32
32 size_t id0 = GetDim(index, 0);
(gdb)
On Intel GT, the 32b IP register holds the offset from the 64b
Instruction Base Address, which we model as virtual 'isabase'
register.
We port async support from linux-low.cc. There is no need to mask
SIGCHLD during pipe initialization as we're not using ptrace. We rely
on GDB explicitly requesting all-stop/non-stop mode to enable async.
For Level-Zero targets, module load and unload events are sent in the
context of the host process. They block a host system call until an
attached debugger acknowledges the event to give it enough time to place
breakpoints before the module may be used.
Accessing the memory requires specifying a debug session handle, which
is available through a thread. For convenience, we introduce
overloaded versions of read_memory and write_memory that take a
`thread_info *` to denote the context in which memory should be
accessed. The existing read_memory and write_memory target ops use
the overloaded versions simply by passing current_thread.
When GDB sends an interrupt, some threads may turn out to be
unavailable. From GDB's PoV, those threads are not running. However,
they may emit events. In all-stop mode, those events are not fetched
until GDB resumes the threads, due to the synchronous communication
model. In non-stop mode, however, the events may be fetched. When
they are sent to GDB, they can cause confusion, in particular when GDB
commits resume requests. To prevent a state mismatch, we hold events
that arise from unavailable threads whose resume state is 'stop'.
Once GDB sends resume requests for these threads, we unleash the
event.
Co-authored-by: Tankut Baris Aktemur <tankut.baris.aktemur@intel.com>
Co-authored-by: Natalia Saiapova <natalia.saiapova@intel.com>
Cc: <config-patches@gnu.org>
Cc: <binutils@sourceware.org>
---
config.sub | 3 +-
gdbserver/Makefile.in | 4 +-
gdbserver/config.in | 6 +
gdbserver/configure | 500 ++++++++
gdbserver/configure.ac | 18 +
gdbserver/configure.srv | 4 +
gdbserver/gdbthread.h | 2 +-
gdbserver/intelgt-ze-low.cc | 1016 +++++++++++++++
gdbserver/ze-low.cc | 2996 +++++++++++++++++++++++++++++++++++++++++++
gdbserver/ze-low.h | 496 +++++++
10 files changed, 5042 insertions(+), 3 deletions(-)
diff --git a/config.sub b/config.sub
index 664ee26124ac817257f076ed1d95f99cd0530fa4..2a54c8a9b410e367c07c38bf061f2bd4be9f74d4 100755
--- a/config.sub
+++ b/config.sub
@@ -2149,6 +2149,7 @@ case $os in
| winnt* \
| xenix* \
| xray* \
+ | ze* \
| zephyr* \
| zvmoe* )
;;
@@ -2211,7 +2212,7 @@ esac
case $kernel-$os-$obj in
linux-gnu*- | linux-android*- | linux-dietlibc*- | linux-llvm*- \
| linux-mlibc*- | linux-musl*- | linux-newlib*- \
- | linux-relibc*- | linux-uclibc*- | linux-ohos*- )
+ | linux-relibc*- | linux-uclibc*- | linux-ohos*- | linux-ze*- )
;;
uclinux-uclibc*- | uclinux-gnu*- )
;;
diff --git a/gdbserver/Makefile.in b/gdbserver/Makefile.in
index dbcc18a85dc70f2aaa8bf71c7d6865cd7f264bd1..230a948b1c02f33b7825ff53e984806ccab7fc3f 100644
--- a/gdbserver/Makefile.in
+++ b/gdbserver/Makefile.in
@@ -104,6 +104,8 @@ LIBIBERTY = $(LIBIBERTY_NORMAL)
GDBSUPPORT_BUILDDIR = ../gdbsupport
GDBSUPPORT = $(GDBSUPPORT_BUILDDIR)/libgdbsupport.a
+LIBZE_LOADER = @LIBZE_LOADER@
+
# gnulib
GNULIB_PARENT_DIR = ..
include $(GNULIB_PARENT_DIR)/gnulib/Makefile.gnulib.inc
@@ -369,7 +371,7 @@ gdbserver$(EXEEXT): $(sort $(OBS)) ${CDEPS} $(LIBGNU) $(LIBIBERTY) \
$(CXXFLAGS) \
-o gdbserver$(EXEEXT) $(OBS) $(GDBSUPPORT) $(LIBGNU) \
$(LIBGNU_EXTRA_LIBS) $(LIBIBERTY) $(INTL) \
- $(GDBSERVER_LIBS) $(XM_CLIBS) $(WIN32APILIBS) $(MAYBE_LIBICONV)
+ $(GDBSERVER_LIBS) $(XM_CLIBS) $(WIN32APILIBS) $(MAYBE_LIBICONV) $(LIBZE_LOADER)
gdbreplay$(EXEEXT): $(sort $(GDBREPLAY_OBS)) $(LIBGNU) $(LIBIBERTY) \
$(INTL_DEPS) $(GDBSUPPORT)
diff --git a/gdbserver/config.in b/gdbserver/config.in
index ead66ce7a5746c5820f06e985332e4df9949b0ba..e4ba29ecd5d3d5f7993d69289bc70bcffb88f6a4 100644
--- a/gdbserver/config.in
+++ b/gdbserver/config.in
@@ -158,9 +158,15 @@
/* Define to 1 if you have the `dl' library (-ldl). */
#undef HAVE_LIBDL
+/* Define if you have the igfxdbg library. */
+#undef HAVE_LIBIGFXDBG
+
/* Define if you have the ipt library. */
#undef HAVE_LIBIPT
+/* Define if you have the ze_loader library. */
+#undef HAVE_LIBZE_LOADER
+
/* Define if you have the xxhash library. */
#undef HAVE_LIBXXHASH
diff --git a/gdbserver/configure b/gdbserver/configure
index b45b55ffde7a23716fd9f788bc7eeecc183e96d6..828105595ca065d291fecdf274ead12b11e1c58a 100755
--- a/gdbserver/configure
+++ b/gdbserver/configure
@@ -631,6 +631,9 @@ srv_xmlfiles
srv_xmlbuiltin
GDBSERVER_LIBS
GDBSERVER_DEPFILES
+LTLIBZE_LOADER
+LIBZE_LOADER
+HAVE_LIBZE_LOADER
RDYNAMIC
REPORT_BUGS_TEXI
REPORT_BUGS_TO
@@ -780,6 +783,8 @@ with_pkgversion
with_bugurl
with_libthread_db
enable_inprocess_agent
+with_libze_loader_prefix
+with_libze_loader_type
'
ac_precious_vars='build_alias
host_alias
@@ -1450,6 +1455,9 @@ Optional Packages:
--with-bugurl=URL Direct users to URL to report a bug
--with-libthread-db=PATH
use given libthread_db directly
+ --with-libze_loader-prefix[=DIR] search for libze_loader in DIR/include and DIR/lib
+ --without-libze_loader-prefix don't search for libze_loader in includedir and libdir
+ --with-libze_loader-type=TYPE type of library to search for (auto/static/shared)
Some influential environment variables:
CC C compiler command
@@ -14725,6 +14733,498 @@ fi
+
+ use_additional=yes
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+
+ eval additional_includedir=\"$includedir\"
+ eval additional_libdir=\"$libdir\"
+
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+
+# Check whether --with-libze_loader-prefix was given.
+if test "${with_libze_loader_prefix+set}" = set; then :
+ withval=$with_libze_loader_prefix;
+ if test "X$withval" = "Xno"; then
+ use_additional=no
+ else
+ if test "X$withval" = "X"; then
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+
+ eval additional_includedir=\"$includedir\"
+ eval additional_libdir=\"$libdir\"
+
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ else
+ additional_includedir="$withval/include"
+ additional_libdir="$withval/lib"
+ fi
+ fi
+
+fi
+
+
+# Check whether --with-libze_loader-type was given.
+if test "${with_libze_loader_type+set}" = set; then :
+ withval=$with_libze_loader_type; with_libze_loader_type=$withval
+else
+ with_libze_loader_type=auto
+fi
+
+ lib_type=`eval echo \$with_libze_loader_type`
+
+ LIBZE_LOADER=
+ LTLIBZE_LOADER=
+ INCZE_LOADER=
+ rpathdirs=
+ ltrpathdirs=
+ names_already_handled=
+ names_next_round='ze_loader '
+ while test -n "$names_next_round"; do
+ names_this_round="$names_next_round"
+ names_next_round=
+ for name in $names_this_round; do
+ already_handled=
+ for n in $names_already_handled; do
+ if test "$n" = "$name"; then
+ already_handled=yes
+ break
+ fi
+ done
+ if test -z "$already_handled"; then
+ names_already_handled="$names_already_handled $name"
+ uppername=`echo "$name" | sed -e 'y|abcdefghijklmnopqrstuvwxyz./-|ABCDEFGHIJKLMNOPQRSTUVWXYZ___|'`
+ eval value=\"\$HAVE_LIB$uppername\"
+ if test -n "$value"; then
+ if test "$value" = yes; then
+ eval value=\"\$LIB$uppername\"
+ test -z "$value" || LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$value"
+ eval value=\"\$LTLIB$uppername\"
+ test -z "$value" || LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }$value"
+ else
+ :
+ fi
+ else
+ found_dir=
+ found_la=
+ found_so=
+ found_a=
+ if test $use_additional = yes; then
+ if test -n "$shlibext" && test -f "$additional_libdir/lib$name.$shlibext" && test x$lib_type != xstatic; then
+ found_dir="$additional_libdir"
+ found_so="$additional_libdir/lib$name.$shlibext"
+ if test -f "$additional_libdir/lib$name.la"; then
+ found_la="$additional_libdir/lib$name.la"
+ fi
+ elif test x$lib_type != xshared; then
+ if test -f "$additional_libdir/lib$name.$libext"; then
+ found_dir="$additional_libdir"
+ found_a="$additional_libdir/lib$name.$libext"
+ if test -f "$additional_libdir/lib$name.la"; then
+ found_la="$additional_libdir/lib$name.la"
+ fi
+ fi
+ fi
+ fi
+ if test "X$found_dir" = "X"; then
+ for x in $LDFLAGS $LTLIBZE_LOADER; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ case "$x" in
+ -L*)
+ dir=`echo "X$x" | sed -e 's/^X-L//'`
+ if test -n "$shlibext" && test -f "$dir/lib$name.$shlibext" && test x$lib_type != xstatic; then
+ found_dir="$dir"
+ found_so="$dir/lib$name.$shlibext"
+ if test -f "$dir/lib$name.la"; then
+ found_la="$dir/lib$name.la"
+ fi
+ elif test x$lib_type != xshared; then
+ if test -f "$dir/lib$name.$libext"; then
+ found_dir="$dir"
+ found_a="$dir/lib$name.$libext"
+ if test -f "$dir/lib$name.la"; then
+ found_la="$dir/lib$name.la"
+ fi
+ fi
+ fi
+ ;;
+ esac
+ if test "X$found_dir" != "X"; then
+ break
+ fi
+ done
+ fi
+ if test "X$found_dir" != "X"; then
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }-L$found_dir -l$name"
+ if test "X$found_so" != "X"; then
+ if test "$enable_rpath" = no || test "X$found_dir" = "X/usr/lib"; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$found_so"
+ else
+ haveit=
+ for x in $ltrpathdirs; do
+ if test "X$x" = "X$found_dir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ ltrpathdirs="$ltrpathdirs $found_dir"
+ fi
+ if test "$hardcode_direct" = yes; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$found_so"
+ else
+ if test -n "$hardcode_libdir_flag_spec" && test "$hardcode_minus_L" = no; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$found_so"
+ haveit=
+ for x in $rpathdirs; do
+ if test "X$x" = "X$found_dir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ rpathdirs="$rpathdirs $found_dir"
+ fi
+ else
+ haveit=
+ for x in $LDFLAGS $LIBZE_LOADER; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ if test "X$x" = "X-L$found_dir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-L$found_dir"
+ fi
+ if test "$hardcode_minus_L" != no; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$found_so"
+ else
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-l$name"
+ fi
+ fi
+ fi
+ fi
+ else
+ if test "X$found_a" != "X"; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$found_a"
+ else
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-L$found_dir -l$name"
+ fi
+ fi
+ additional_includedir=
+ case "$found_dir" in
+ */lib | */lib/)
+ basedir=`echo "X$found_dir" | sed -e 's,^X,,' -e 's,/lib/*$,,'`
+ additional_includedir="$basedir/include"
+ ;;
+ esac
+ if test "X$additional_includedir" != "X"; then
+ if test "X$additional_includedir" != "X/usr/include"; then
+ haveit=
+ if test "X$additional_includedir" = "X/usr/local/include"; then
+ if test -n "$GCC"; then
+ case $host_os in
+ linux*) haveit=yes;;
+ esac
+ fi
+ fi
+ if test -z "$haveit"; then
+ for x in $CPPFLAGS $INCZE_LOADER; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ if test "X$x" = "X-I$additional_includedir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ if test -d "$additional_includedir"; then
+ INCZE_LOADER="${INCZE_LOADER}${INCZE_LOADER:+ }-I$additional_includedir"
+ fi
+ fi
+ fi
+ fi
+ fi
+ if test -n "$found_la"; then
+ save_libdir="$libdir"
+ case "$found_la" in
+ */* | *\\*) . "$found_la" ;;
+ *) . "./$found_la" ;;
+ esac
+ libdir="$save_libdir"
+ for dep in $dependency_libs; do
+ case "$dep" in
+ -L*)
+ additional_libdir=`echo "X$dep" | sed -e 's/^X-L//'`
+ if test "X$additional_libdir" != "X/usr/lib"; then
+ haveit=
+ if test "X$additional_libdir" = "X/usr/local/lib"; then
+ if test -n "$GCC"; then
+ case $host_os in
+ linux*) haveit=yes;;
+ esac
+ fi
+ fi
+ if test -z "$haveit"; then
+ haveit=
+ for x in $LDFLAGS $LIBZE_LOADER; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ if test "X$x" = "X-L$additional_libdir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ if test -d "$additional_libdir"; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-L$additional_libdir"
+ fi
+ fi
+ haveit=
+ for x in $LDFLAGS $LTLIBZE_LOADER; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ if test "X$x" = "X-L$additional_libdir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ if test -d "$additional_libdir"; then
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }-L$additional_libdir"
+ fi
+ fi
+ fi
+ fi
+ ;;
+ -R*)
+ dir=`echo "X$dep" | sed -e 's/^X-R//'`
+ if test "$enable_rpath" != no; then
+ haveit=
+ for x in $rpathdirs; do
+ if test "X$x" = "X$dir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ rpathdirs="$rpathdirs $dir"
+ fi
+ haveit=
+ for x in $ltrpathdirs; do
+ if test "X$x" = "X$dir"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ ltrpathdirs="$ltrpathdirs $dir"
+ fi
+ fi
+ ;;
+ -l*)
+ names_next_round="$names_next_round "`echo "X$dep" | sed -e 's/^X-l//'`
+ ;;
+ *.la)
+ names_next_round="$names_next_round "`echo "X$dep" | sed -e 's,^X.*/,,' -e 's,^lib,,' -e 's,\.la$,,'`
+ ;;
+ *)
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$dep"
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }$dep"
+ ;;
+ esac
+ done
+ fi
+ else
+ if test "x$lib_type" = "xauto" || test "x$lib_type" = "xshared"; then
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-l$name"
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }-l$name"
+ else
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }-l:lib$name.$libext"
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }-l:lib$name.$libext"
+ fi
+ fi
+ fi
+ fi
+ done
+ done
+ if test "X$rpathdirs" != "X"; then
+ if test -n "$hardcode_libdir_separator"; then
+ alldirs=
+ for found_dir in $rpathdirs; do
+ alldirs="${alldirs}${alldirs:+$hardcode_libdir_separator}$found_dir"
+ done
+ acl_save_libdir="$libdir"
+ libdir="$alldirs"
+ eval flag=\"$hardcode_libdir_flag_spec\"
+ libdir="$acl_save_libdir"
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$flag"
+ else
+ for found_dir in $rpathdirs; do
+ acl_save_libdir="$libdir"
+ libdir="$found_dir"
+ eval flag=\"$hardcode_libdir_flag_spec\"
+ libdir="$acl_save_libdir"
+ LIBZE_LOADER="${LIBZE_LOADER}${LIBZE_LOADER:+ }$flag"
+ done
+ fi
+ fi
+ if test "X$ltrpathdirs" != "X"; then
+ for found_dir in $ltrpathdirs; do
+ LTLIBZE_LOADER="${LTLIBZE_LOADER}${LTLIBZE_LOADER:+ }-R$found_dir"
+ done
+ fi
+
+
+ ac_save_CPPFLAGS="$CPPFLAGS"
+
+ for element in $INCZE_LOADER; do
+ haveit=
+ for x in $CPPFLAGS; do
+
+ acl_save_prefix="$prefix"
+ prefix="$acl_final_prefix"
+ acl_save_exec_prefix="$exec_prefix"
+ exec_prefix="$acl_final_exec_prefix"
+ eval x=\"$x\"
+ exec_prefix="$acl_save_exec_prefix"
+ prefix="$acl_save_prefix"
+
+ if test "X$x" = "X$element"; then
+ haveit=yes
+ break
+ fi
+ done
+ if test -z "$haveit"; then
+ CPPFLAGS="${CPPFLAGS}${CPPFLAGS:+ }$element"
+ fi
+ done
+
+
+ { $as_echo "$as_me:${as_lineno-$LINENO}: checking for libze_loader" >&5
+$as_echo_n "checking for libze_loader... " >&6; }
+if ${ac_cv_libze_loader+:} false; then :
+ $as_echo_n "(cached) " >&6
+else
+
+ ac_save_LIBS="$LIBS"
+ LIBS="$LIBS $LIBZE_LOADER"
+ cat confdefs.h - <<_ACEOF >conftest.$ac_ext
+/* end confdefs.h. */
+#include "level_zero/zet_api.h"
+int
+main ()
+{
+zeInit (0);
+ ;
+ return 0;
+}
+_ACEOF
+if ac_fn_c_try_link "$LINENO"; then :
+ ac_cv_libze_loader=yes
+else
+ ac_cv_libze_loader=no
+fi
+rm -f core conftest.err conftest.$ac_objext \
+ conftest$ac_exeext conftest.$ac_ext
+ LIBS="$ac_save_LIBS"
+
+fi
+{ $as_echo "$as_me:${as_lineno-$LINENO}: result: $ac_cv_libze_loader" >&5
+$as_echo "$ac_cv_libze_loader" >&6; }
+ if test "$ac_cv_libze_loader" = yes; then
+ HAVE_LIBZE_LOADER=yes
+
+$as_echo "#define HAVE_LIBZE_LOADER 1" >>confdefs.h
+
+ { $as_echo "$as_me:${as_lineno-$LINENO}: checking how to link with libze_loader" >&5
+$as_echo_n "checking how to link with libze_loader... " >&6; }
+ { $as_echo "$as_me:${as_lineno-$LINENO}: result: $LIBZE_LOADER" >&5
+$as_echo "$LIBZE_LOADER" >&6; }
+ else
+ HAVE_LIBZE_LOADER=no
+ CPPFLAGS="$ac_save_CPPFLAGS"
+ LIBZE_LOADER=
+ LTLIBZE_LOADER=
+ fi
+
+
+
+
+
+
+
+case "${target}" in
+ intelgt-*-ze)
+ if test "$HAVE_LIBZE_LOADER" != yes; then
+ as_fn_error $? "libze_loader is missing or unusable" "$LINENO" 5
+ fi
+ ;;
+ *)
+ # Do not link libze_loader spuriously
+ HAVE_LIBZE_LOADER=no
+ LIBZE_LOADER=
+ LTLIBZE_LOADER=
+ ;;
+esac
+
+
+
+
+
+
+
+
GNULIB=../gnulib/import
GNULIB_STDINT_H=
diff --git a/gdbserver/configure.ac b/gdbserver/configure.ac
index bd2cac8d35db87915b7c20e5333b191caf7a485f..ce07a4d470598bf3623435c2b8f71478e884f997 100644
--- a/gdbserver/configure.ac
+++ b/gdbserver/configure.ac
@@ -403,6 +403,24 @@ if $want_ipa ; then
fi
fi
+dnl check for the ze loader
+AC_LIB_HAVE_LINKFLAGS([ze_loader], [], [#include "level_zero/zet_api.h"],
+ [zeInit (0);],)
+
+case "${target}" in
+ intelgt-*-ze)
+ if test "$HAVE_LIBZE_LOADER" != yes; then
+ AC_MSG_ERROR([libze_loader is missing or unusable])
+ fi
+ ;;
+ *)
+ # Do not link libze_loader spuriously
+ HAVE_LIBZE_LOADER=no
+ LIBZE_LOADER=
+ LTLIBZE_LOADER=
+ ;;
+esac
+
AC_SUBST(GDBSERVER_DEPFILES)
AC_SUBST(GDBSERVER_LIBS)
AC_SUBST(srv_xmlbuiltin)
diff --git a/gdbserver/configure.srv b/gdbserver/configure.srv
index 9a6944b6e4eb570d9905c8dd76e3ee3bb078bbf8..1c54bb8b992003726c34fc1db3f89b3df3a0dde4 100644
--- a/gdbserver/configure.srv
+++ b/gdbserver/configure.srv
@@ -147,6 +147,10 @@ case "${gdbserver_host}" in
srv_tgtobj="$srv_linux_obj linux-ia64-low.o"
srv_linux_usrregs=yes
;;
+ intelgt-*-ze) srv_regobj=""
+ srv_xmlfiles=""
+ srv_tgtobj="ze-low.o intelgt-ze-low.o arch/intelgt.o"
+ ;;
loongarch*-*-linux*) srv_tgtobj="arch/loongarch.o linux-loongarch-low.o"
srv_tgtobj="${srv_tgtobj} ${srv_linux_obj}"
srv_tgtobj="$srv_tgtobj nat/loongarch-hw-point.o"
diff --git a/gdbserver/gdbthread.h b/gdbserver/gdbthread.h
index 5e6130106f5f44a275f9e9873c72b8bfc0cba4d8..7f280843c301baa27d8c30709ff8a75a039f63e9 100644
--- a/gdbserver/gdbthread.h
+++ b/gdbserver/gdbthread.h
@@ -42,7 +42,7 @@ struct thread_info : public intrusive_list_node<thread_info>
void set_regcache (std::unique_ptr<struct regcache> regcache)
{ m_regcache = std::move (regcache); }
- void *target_data ()
+ void *target_data () const
{ return m_target_data; }
/* The id of this thread. */
diff --git a/gdbserver/intelgt-ze-low.cc b/gdbserver/intelgt-ze-low.cc
new file mode 100644
index 0000000000000000000000000000000000000000..ac0c157a46a778f6014f61192b3bd6fa7c8c94dc
--- /dev/null
+++ b/gdbserver/intelgt-ze-low.cc
@@ -0,0 +1,1016 @@
+/* Target interface for Intel GT based on Level-Zero for gdbserver.
+
+ Copyright (C) 2020-2025 Free Software Foundation, Inc.
+
+ This file is part of GDB.
+
+ 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 "ze-low.h"
+#include "arch/intelgt.h"
+#include "gdbsupport/osabi.h"
+
+#include <level_zero/zet_intel_gpu_debug.h>
+#include <iomanip>
+#include <sstream>
+
+
+/* FIXME make into a target method? */
+int using_threads = 1;
+
+/* Convenience macros. */
+
+#define dprintf(...) \
+ do \
+ { \
+ if (debug_threads) \
+ { \
+ debug_printf ("%s: ", __FUNCTION__); \
+ debug_printf (__VA_ARGS__); \
+ debug_printf ("\n"); \
+ } \
+ } \
+ while (0)
+
+
+/* Determine the most suitable type to be used for a register with bit size
+ BITSIZE and element size ELEMSIZE. */
+
+static const char *
+intelgt_uint_reg_type (tdesc_feature *feature, uint32_t bitsize,
+ uint32_t elemsize)
+{
+ if (0 != (bitsize % elemsize))
+ error (_("unsupported combination of bitsize %" PRIu32 "and elemsize %"
+ PRIu32), bitsize, elemsize);
+ if ((elemsize < 8) || (elemsize > 128) || ((elemsize & (elemsize - 1)) != 0))
+ error (_("unsupported elemsize %" PRIu32), elemsize);
+
+ char type_name[20];
+ snprintf (type_name, sizeof (type_name), "uint%u", elemsize);
+ tdesc_type *type = tdesc_named_type (feature, type_name);
+
+ if (elemsize == bitsize)
+ return type->name.c_str ();
+
+ uint32_t elements = bitsize / elemsize;
+ snprintf (type_name, sizeof (type_name), "vector%ux%u", elements,
+ elemsize);
+ tdesc_type *vector
+ = tdesc_create_vector (feature, type_name, type, elements);
+
+ return vector->name.c_str ();
+}
+
+/* Add a (uniform) register set to FEATURE. */
+
+static void
+intelgt_add_regset (tdesc_feature *feature, long ®num,
+ const char *prefix, uint32_t count, const char *group,
+ uint32_t bitsize, const char *type, expedite_t &expedite)
+{
+ for (uint32_t reg = 0; reg < count; ++reg)
+ {
+ std::string name = std::string (prefix) + std::to_string (reg);
+
+ tdesc_create_reg (feature, name.c_str (), regnum++, 1, group,
+ bitsize, type);
+ }
+}
+
+/* Control Register details. */
+
+enum
+{
+ /* The position of the Breakpoint Suppress bit in CR0.0. */
+ INTELGT_CR0_0_BREAKPOINT_SUPPRESS = 15,
+
+ /* The position of the Breakpoint Status and Control bit in CR0.1. */
+ INTELGT_CR0_1_BREAKPOINT_STATUS = 31,
+
+ /* The position of the External Halt Status and Control bit in CR0.1. */
+ INTELGT_CR0_1_EXTERNAL_HALT_STATUS = 30,
+
+ /* The position of the Software Exception Control bit in CR0.1. */
+ INTELGT_CR0_1_SOFTWARE_EXCEPTION_CONTROL = 29,
+
+ /* The position of the Illegal Opcode Exception Status bit in CR0.1. */
+ INTELGT_CR0_1_ILLEGAL_OPCODE_STATUS = 28,
+
+ /* The position of the Force Exception Status and Control bit in CR0.1. */
+ INTELGT_CR0_1_FORCE_EXCEPTION_STATUS = 26,
+
+ /* The position of the Page Fault Status bit in CR0.1.
+ This is a software convention using a reserved bit to indicate
+ page faults by the user mode driver. */
+ INTELGT_CR0_1_PAGEFAULT_STATUS = 16,
+};
+
+/* Return CR0.SUBREG in REGCACHE. */
+
+static uint32_t
+intelgt_read_cr0 (regcache *regcache, int subreg)
+{
+ int cr0regno = find_regno (regcache->tdesc, "cr0");
+ int cr0size = register_size (regcache->tdesc, cr0regno);
+ uint32_t cr0[16];
+ gdb_assert (cr0size <= sizeof (cr0));
+ gdb_assert (cr0size >= sizeof (cr0[0]) * (subreg + 1));
+ collect_register (regcache, cr0regno, cr0);
+
+ enum register_status cr0status = regcache->get_register_status (cr0regno);
+ switch (cr0status)
+ {
+ case REG_VALID:
+ return cr0[subreg];
+
+ case REG_UNKNOWN:
+ internal_error (_("unknown register 'cr0'."));
+
+ case REG_UNAVAILABLE:
+ error (_("cr0 is not available"));
+ }
+
+ internal_error (_("unknown register status: %d."), cr0status);
+}
+
+/* Write VALUE into CR0.SUBREG in REGCACHE. */
+
+static void
+intelgt_write_cr0 (regcache *regcache, int subreg, uint32_t value)
+{
+ int cr0regno = find_regno (regcache->tdesc, "cr0");
+ int cr0size = register_size (regcache->tdesc, cr0regno);
+ uint32_t cr0[16];
+ gdb_assert (cr0size <= sizeof (cr0));
+ gdb_assert (cr0size >= sizeof (cr0[0]) * (subreg + 1));
+ collect_register (regcache, cr0regno, cr0);
+
+ enum register_status cr0status = regcache->get_register_status (cr0regno);
+ switch (cr0status)
+ {
+ case REG_VALID:
+ cr0[subreg] = value;
+ supply_register (regcache, cr0regno, cr0);
+ return;
+
+ case REG_UNKNOWN:
+ internal_error (_("unknown register 'cr0'."));
+
+ case REG_UNAVAILABLE:
+ error (_("cr0 is not available"));
+ }
+
+ internal_error (_("unknown register status: %d."), cr0status);
+}
+
+/* Return CR0.SUBREG for TP. */
+
+static uint32_t
+intelgt_read_cr0 (thread_info *tp, int subreg)
+{
+ struct regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ return intelgt_read_cr0 (regcache, subreg);
+}
+
+/* Write VALUE into CR0.SUBREG for TP. */
+
+static void
+intelgt_write_cr0 (thread_info *tp, int subreg, uint32_t value)
+{
+ struct regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ intelgt_write_cr0 (regcache, subreg, value);
+}
+
+/* Return a human-readable device UUID string. */
+
+static std::string
+device_uuid_str (const uint8_t uuid[], size_t size)
+{
+ std::stringstream sstream;
+ for (int i = size - 1; i >= 0; --i)
+ sstream << std::hex << std::setfill ('0') << std::setw (2)
+ << static_cast<int> (uuid[i]);
+
+ return sstream.str ();
+}
+
+static uint32_t
+get_device_id (ze_device_info *device)
+{
+ gdb_assert (device != nullptr);
+ return device->properties.deviceId;
+}
+
+/* Target op definitions for Intel GT target based on Level-Zero. */
+
+class intelgt_ze_target : public ze_target
+{
+public:
+ const gdb_byte *sw_breakpoint_from_kind (int kind, int *size) override;
+
+ bool supports_stopped_by_sw_breakpoint () override { return true; }
+ bool stopped_by_sw_breakpoint () override;
+
+ CORE_ADDR read_pc (regcache *regcache) override;
+ void write_pc (regcache *regcache, CORE_ADDR pc) override;
+
+protected:
+ bool is_device_supported
+ (const ze_device_properties_t &,
+ const std::vector<zet_debug_regset_properties_t> &) override;
+
+ target_desc *create_tdesc
+ (ze_device_info *dinfo,
+ const std::vector<zet_debug_regset_properties_t> &,
+ const ze_pci_ext_properties_t &) override;
+
+ target_stop_reason get_stop_reason (thread_info *, gdb_signal &) override;
+
+ void prepare_thread_resume (thread_info *tp) override;
+
+ /* Read one instruction from memory at PC into BUFFER and return the
+ number of bytes read on success or a negative errno error code.
+
+ BUFFER must be intelgt::MAX_INST_LENGTH bytes long. */
+ int read_inst (thread_info *tp, CORE_ADDR pc,
+ gdb::array_view<gdb_byte> buffer);
+
+ bool is_at_breakpoint (thread_info *tp) override;
+ bool is_at_eot (thread_info *tp);
+
+ bool erratum_18020355813 (thread_info *tp);
+
+private:
+ /* Add a register set for REGPROP on DEVICE to REGSETS and increment REGNUM
+ accordingly.
+
+ May optionally add registers to EXPEDITE. */
+ void add_regset (target_desc *tdesc, const ze_device_info &dinfo,
+ const zet_debug_regset_properties_t ®prop,
+ long ®num, ze_regset_info_t ®sets,
+ expedite_t &expedite);
+};
+
+const gdb_byte *
+intelgt_ze_target::sw_breakpoint_from_kind (int kind, int *size)
+{
+ /* We do not support breakpoint instructions.
+
+ Use gdbarch methods that use read/write memory target operations for
+ setting s/w breakopints. */
+ *size = 0;
+ return nullptr;
+}
+
+bool
+intelgt_ze_target::stopped_by_sw_breakpoint ()
+{
+ const ze_thread_info *zetp = ze_thread (current_thread);
+ if (zetp == nullptr)
+ return false;
+
+ ptid_t ptid = current_thread->id;
+
+ if (zetp->exec_state != ZE_THREAD_STATE_STOPPED)
+ {
+ dprintf ("not-stopped thread %s", ptid.to_string ().c_str ());
+ return false;
+ }
+
+ return (zetp->stop_reason == TARGET_STOPPED_BY_SW_BREAKPOINT);
+}
+
+CORE_ADDR
+intelgt_ze_target::read_pc (regcache *regcache)
+{
+ uint32_t ip = intelgt_read_cr0 (regcache, 2);
+ uint64_t isabase;
+ collect_register_by_name (regcache, "isabase", &isabase);
+
+ if (UINT32_MAX < ip)
+ warning (_("IP '0x%" PRIx32 "' outside of ISA range."), ip);
+
+ CORE_ADDR pc = (CORE_ADDR) isabase + (CORE_ADDR) ip;
+ if (pc < isabase)
+ warning (_("PC '%s' outside of ISA range."),
+ core_addr_to_string_nz (pc));
+
+ return pc;
+}
+
+void
+intelgt_ze_target::write_pc (regcache *regcache, CORE_ADDR pc)
+{
+ uint64_t isabase;
+ collect_register_by_name (regcache, "isabase", &isabase);
+
+ if (pc < isabase)
+ error (_("PC '%s' outside of ISA range."), core_addr_to_string_nz (pc));
+
+ pc -= isabase;
+ if (UINT32_MAX < pc)
+ error (_("PC '%s' outside of ISA range."), core_addr_to_string_nz (pc));
+
+ intelgt_write_cr0 (regcache, 2, (uint32_t) pc);
+}
+
+bool
+intelgt_ze_target::is_device_supported
+ (const ze_device_properties_t &properties,
+ const std::vector<zet_debug_regset_properties_t> ®set_properties)
+{
+ if (properties.type != ZE_DEVICE_TYPE_GPU)
+ {
+ dprintf ("non-gpu (%x) device (%" PRIx32 "): %s", properties.type,
+ properties.deviceId, properties.name);
+ return false;
+ }
+
+ if (properties.vendorId != 0x8086)
+ {
+ dprintf ("unknown vendor (%" PRIx32 ") of device (%" PRIx32 "): %s",
+ properties.vendorId, properties.deviceId, properties.name);
+ return false;
+ }
+
+ /* We need a few registers to support an Intel GT device.
+
+ Those are registers that GDB itself uses. Without those, we might run into
+ internal errors at some point. We need others, too, that may be referenced
+ in debug information. */
+ bool have_grf = false;
+ bool have_isabase = false;
+ bool have_cr = false;
+ bool have_sr = false;
+ bool have_ce = false;
+ for (const zet_debug_regset_properties_t ®prop : regset_properties)
+ {
+ if (regprop.count < 1)
+ {
+ warning (_("Ignoring empty regset %u in %s."), regprop.type,
+ properties.name);
+ continue;
+ }
+
+ switch (regprop.type)
+ {
+ case ZET_DEBUG_REGSET_TYPE_GRF_INTEL_GPU:
+ have_grf = true;
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_CE_INTEL_GPU:
+ have_ce = true;
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_CR_INTEL_GPU:
+ have_cr = true;
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_SR_INTEL_GPU:
+ have_sr = true;
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_SBA_INTEL_GPU:
+ /* We need 'isabase', which is at position 5 in version 1. */
+ if ((regprop.version == 0) && (regprop.count >= 5))
+ have_isabase = true;
+ else
+ warning (_("Ignoring unknown SBA regset version %u in %s."),
+ regprop.version, properties.name);
+ break;
+ }
+ }
+
+ if (have_grf && have_isabase && have_cr && have_sr && have_ce)
+ return true;
+
+ dprintf ("unsupported device (%" PRIx32 "): %s", properties.deviceId,
+ properties.name);
+ return false;
+}
+
+target_desc *
+intelgt_ze_target::create_tdesc
+ (ze_device_info *dinfo,
+ const std::vector<zet_debug_regset_properties_t> ®set_properties,
+ const ze_pci_ext_properties_t &pci_properties)
+{
+ const ze_device_properties_t &properties = dinfo->properties;
+
+ if (properties.vendorId != 0x8086)
+ error (_("unknown vendor (%" PRIx32 ") of device (%" PRIx32 "): %s"),
+ properties.vendorId, properties.deviceId, properties.name);
+
+ target_desc_up tdesc = allocate_target_description ();
+ set_tdesc_architecture (tdesc.get (), "intelgt");
+ set_tdesc_osabi (tdesc.get (), GDB_OSABI_LINUX);
+
+ std::string device_uuid = device_uuid_str (
+ dinfo->properties.uuid.id, sizeof (dinfo->properties.uuid.id));
+ const uint32_t total_cores = (properties.numSlices
+ * properties.numSubslicesPerSlice
+ * properties.numEUsPerSubslice);
+ const uint32_t total_threads = (total_cores * properties.numThreadsPerEU);
+
+ tdesc_device *device_info = new tdesc_device ();
+ device_info->vendor_id = properties.vendorId;
+ device_info->target_id = properties.deviceId;
+ device_info->name = properties.name;
+ device_info->pci_slot = string_printf ("%02" PRIx32 ":%02" PRIx32
+ ".%" PRId32,
+ pci_properties.address.bus,
+ pci_properties.address.device,
+ pci_properties.address.function);
+ device_info->uuid = device_uuid;
+ device_info->total_cores = total_cores;
+ device_info->total_threads = total_threads;
+
+ if (properties.flags & ZE_DEVICE_PROPERTY_FLAG_SUBDEVICE)
+ device_info->subdevice_id = properties.subdeviceId;
+
+ set_tdesc_device_info (tdesc.get (), device_info);
+
+ long regnum = 0;
+ for (const zet_debug_regset_properties_t ®prop : regset_properties)
+ add_regset (tdesc.get (), *dinfo, regprop, regnum,
+ dinfo->regsets, dinfo->expedite);
+
+ /* Tdesc expects a nullptr-terminated array. */
+ dinfo->expedite.push_back (nullptr);
+
+ init_target_desc (tdesc.get (), dinfo->expedite.data (), GDB_OSABI_LINUX);
+ return tdesc.release ();
+}
+
+target_stop_reason
+intelgt_ze_target::get_stop_reason (thread_info *tp, gdb_signal &signal)
+{
+ ze_device_thread_t thread = ze_thread_id (tp);
+ uint32_t cr0[3] = {
+ intelgt_read_cr0 (tp, 0),
+ intelgt_read_cr0 (tp, 1),
+ intelgt_read_cr0 (tp, 2)
+ };
+
+ dprintf ("thread %s (%s) stopped, cr0.0=%" PRIx32 ", .1=%" PRIx32
+ " [ %s%s%s%s%s%s], .2=%" PRIx32 ".", tp->id.to_string ().c_str (),
+ ze_thread_id_str (thread).c_str (), cr0[0], cr0[1],
+ (((cr0[1] & (1 << INTELGT_CR0_1_BREAKPOINT_STATUS)) != 0)
+ ? "bp " : ""),
+ (((cr0[1] & (1 << INTELGT_CR0_1_ILLEGAL_OPCODE_STATUS)) != 0)
+ ? "ill " : ""),
+ (((cr0[1] & (1 << INTELGT_CR0_1_FORCE_EXCEPTION_STATUS)) != 0)
+ ? "fe " : ""),
+ (((cr0[1] & (1 << INTELGT_CR0_1_SOFTWARE_EXCEPTION_CONTROL)) != 0)
+ ? "sw " : ""),
+ (((cr0[1] & (1 << INTELGT_CR0_1_EXTERNAL_HALT_STATUS)) != 0)
+ ? "eh " : ""),
+ (((cr0[1] & (1 << INTELGT_CR0_1_PAGEFAULT_STATUS)) != 0)
+ ? "pf " : ""),
+ cr0[2]);
+
+ if ((cr0[1] & (1 << INTELGT_CR0_1_PAGEFAULT_STATUS)) != 0)
+ {
+ cr0[1] &= ~(1 << INTELGT_CR0_1_PAGEFAULT_STATUS);
+ intelgt_write_cr0 (tp, 1, cr0[1]);
+
+ signal = GDB_SIGNAL_SEGV;
+ return TARGET_STOPPED_BY_NO_REASON;
+ }
+
+ if ((cr0[1] & (1 << INTELGT_CR0_1_BREAKPOINT_STATUS)) != 0)
+ {
+ cr0[1] &= ~(1 << INTELGT_CR0_1_BREAKPOINT_STATUS);
+ intelgt_write_cr0 (tp, 1, cr0[1]);
+
+ /* We cannot distinguish a single step exception from a breakpoint
+ exception just by looking at CR0.
+
+ We could inspect the instruction to see if the breakpoint bit is
+ set. Or we could check the resume type and assume that we set
+ things up correctly for single-stepping before we resumed. */
+ const ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ switch (zetp->resume_state)
+ {
+ case ZE_THREAD_RESUME_STEP:
+ signal = GDB_SIGNAL_TRAP;
+ return TARGET_STOPPED_BY_SINGLE_STEP;
+
+ case ZE_THREAD_RESUME_RUN:
+ case ZE_THREAD_RESUME_NONE:
+ /* On some devices, we may get spurious breakpoint exceptions. */
+ if (erratum_18020355813 (tp))
+ {
+ ze_device_thread_t zeid = ze_thread_id (tp);
+
+ dprintf ("applying #18020355813 workaround for thread "
+ "%s (%s)", tp->id.to_string ().c_str (),
+ ze_thread_id_str (zeid).c_str ());
+
+ signal = GDB_SIGNAL_0;
+ return TARGET_STOPPED_BY_NO_REASON;
+ }
+
+ [[fallthrough]];
+ case ZE_THREAD_RESUME_STOP:
+ signal = GDB_SIGNAL_TRAP;
+ return TARGET_STOPPED_BY_SW_BREAKPOINT;
+ }
+ }
+
+ if ((cr0[1] & (1 << INTELGT_CR0_1_ILLEGAL_OPCODE_STATUS)) != 0)
+ {
+ cr0[1] &= ~(1 << INTELGT_CR0_1_ILLEGAL_OPCODE_STATUS);
+ intelgt_write_cr0 (tp, 1, cr0[1]);
+
+ signal = GDB_SIGNAL_ILL;
+ return TARGET_STOPPED_BY_NO_REASON;
+ }
+
+ if ((cr0[1] & (1 << INTELGT_CR0_1_SOFTWARE_EXCEPTION_CONTROL)) != 0)
+ {
+ cr0[1] &= ~(1 << INTELGT_CR0_1_SOFTWARE_EXCEPTION_CONTROL);
+ intelgt_write_cr0 (tp, 1, cr0[1]);
+
+ signal = GDB_EXC_SOFTWARE;
+ return TARGET_STOPPED_BY_NO_REASON;
+ }
+
+ if ((cr0[1] & ((1 << INTELGT_CR0_1_FORCE_EXCEPTION_STATUS)
+ | (1 << INTELGT_CR0_1_EXTERNAL_HALT_STATUS))) != 0)
+ {
+ cr0[1] &= ~(1 << INTELGT_CR0_1_FORCE_EXCEPTION_STATUS);
+ cr0[1] &= ~(1 << INTELGT_CR0_1_EXTERNAL_HALT_STATUS);
+ intelgt_write_cr0 (tp, 1, cr0[1]);
+
+ signal = GDB_SIGNAL_TRAP;
+ return TARGET_STOPPED_BY_NO_REASON;
+ }
+
+ signal = GDB_SIGNAL_UNKNOWN;
+ return TARGET_STOPPED_BY_NO_REASON;
+}
+
+int
+intelgt_ze_target::read_inst (thread_info *tp, CORE_ADDR pc,
+ gdb::array_view<gdb_byte> buffer)
+{
+ gdb_assert (buffer.size () >= intelgt::MAX_INST_LENGTH);
+
+ int status = read_memory (tp, pc, buffer.data (), intelgt::MAX_INST_LENGTH);
+ if (status == 0)
+ return intelgt::MAX_INST_LENGTH;
+
+ status = read_memory (tp, pc, buffer.data (), intelgt::COMPACT_INST_LENGTH);
+ if (status > 0)
+ return status;
+
+ uint32_t device_id = get_device_id (ze_thread_device (tp));
+ if (intelgt::inst_length (buffer, device_id) == intelgt::MAX_INST_LENGTH)
+ return -EIO;
+
+ memset (buffer.begin () + intelgt::COMPACT_INST_LENGTH, 0,
+ intelgt::MAX_INST_LENGTH - intelgt::COMPACT_INST_LENGTH);
+
+ return intelgt::COMPACT_INST_LENGTH;
+}
+
+bool
+intelgt_ze_target::is_at_breakpoint (thread_info *tp)
+{
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ CORE_ADDR pc = read_pc (regcache);
+
+ gdb_byte inst[intelgt::MAX_INST_LENGTH];
+ int status = read_inst (tp, pc, inst);
+ if (status < 0)
+ return false;
+
+ uint32_t device_id = get_device_id (ze_thread_device (tp));
+ return intelgt::has_breakpoint (inst, device_id);
+}
+
+bool
+intelgt_ze_target::is_at_eot (thread_info *tp)
+{
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ CORE_ADDR pc = read_pc (regcache);
+
+ gdb_byte inst[intelgt::MAX_INST_LENGTH];
+ int status = read_inst (tp, pc, inst);
+ if (status < 0)
+ {
+ ze_device_thread_t zeid = ze_thread_id (tp);
+
+ warning (_("error reading memory for thread %s (%s) at 0x%"
+ PRIx64), tp->id.to_string ().c_str (),
+ ze_thread_id_str (zeid).c_str (), pc);
+ return false;
+ }
+
+ uint32_t device_id = get_device_id (ze_thread_device (tp));
+ intelgt::xe_version device_version = intelgt::get_xe_version (device_id);
+ switch (device_version)
+ {
+ case intelgt::XE_HP:
+ case intelgt::XE_HPG:
+ case intelgt::XE_HPC:
+ case intelgt::XE2:
+ case intelgt::XE3:
+ {
+ /* The opcode mask for bits 6:0. */
+ constexpr uint8_t OPC_MASK = 0x7f;
+ switch (inst[0] & OPC_MASK)
+ {
+ case 0x31: /* send */
+ case 0x32: /* sendc */
+ {
+ /* The End Of Thread control. Only used for SEND and
+ SENDC. */
+ constexpr uint8_t CTRL_EOT_SEND = 34;
+ return intelgt::get_inst_bit (inst, CTRL_EOT_SEND);
+ }
+
+ default:
+ return false;
+ }
+ }
+
+ default:
+ error (_("Unsupported device id 0x%" PRIx32), device_id);
+ }
+}
+
+/* Return whether erratum #18020355813 applies. */
+
+bool
+intelgt_ze_target::erratum_18020355813 (thread_info *tp)
+{
+ ze_device_info *device = ze_thread_device (tp);
+
+ /* We may not have a device if we got detached. */
+ if (device == nullptr)
+ return false;
+
+ /* The erratum only applies to Intel devices. */
+ if (device->properties.vendorId != 0x8086)
+ return false;
+
+ uint32_t device_id = get_device_id (device);
+
+ /* The erratum only applies to a range of devices. */
+ switch (intelgt::get_xe_version (device_id))
+ {
+ case intelgt::XE_HPG:
+ case intelgt::XE_HPC:
+ break;
+
+ default:
+ return false;
+ }
+
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ CORE_ADDR pc = read_pc (regcache);
+
+ gdb_byte inst[intelgt::MAX_INST_LENGTH];
+ int status = read_inst (tp, pc, inst);
+ if (status < 0)
+ {
+ ze_device_thread_t zeid = ze_thread_id (tp);
+
+ warning (_("error reading memory for thread %s (%s) at 0x%"
+ PRIx64), tp->id.to_string ().c_str (),
+ ze_thread_id_str (zeid).c_str (), pc);
+ return false;
+ }
+
+ /* The erratum applies to instructions without breakpoint control. */
+ return !intelgt::has_breakpoint (inst, device_id);
+}
+
+void
+intelgt_ze_target::prepare_thread_resume (thread_info *tp)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ uint32_t cr0[3] = {
+ intelgt_read_cr0 (regcache, 0),
+ intelgt_read_cr0 (regcache, 1),
+ intelgt_read_cr0 (regcache, 2)
+ };
+
+ /* The thread is running. We may need to overwrite this below. */
+ zetp->exec_state = ZE_THREAD_STATE_RUNNING;
+
+ /* Clear any potential interrupt indication.
+
+ We leave other exception indications so the exception would be
+ reported again and can be handled by GDB. */
+ cr0[1] &= ~(1 << INTELGT_CR0_1_FORCE_EXCEPTION_STATUS);
+ cr0[1] &= ~(1 << INTELGT_CR0_1_EXTERNAL_HALT_STATUS);
+
+ /* Distinguish stepping and continuing. */
+ switch (zetp->resume_state)
+ {
+ case ZE_THREAD_RESUME_STEP:
+ /* We step by indicating a breakpoint exception, which will be
+ considered on the next instruction.
+
+ This does not work for EOT, though. */
+ if (!is_at_eot (tp))
+ {
+ cr0[0] |= (1 << INTELGT_CR0_0_BREAKPOINT_SUPPRESS);
+ cr0[1] |= (1 << INTELGT_CR0_1_BREAKPOINT_STATUS);
+ break;
+ }
+
+ /* At EOT, the thread dispatch ends and the thread becomes idle.
+
+ There's no point in requesting a single-step exception but we
+ need to inject an event to tell GDB that the step completed. */
+ zetp->exec_state = ZE_THREAD_STATE_UNAVAILABLE;
+ zetp->waitstatus.set_unavailable ();
+
+ [[fallthrough]];
+ case ZE_THREAD_RESUME_RUN:
+ cr0[1] &= ~(1 << INTELGT_CR0_1_BREAKPOINT_STATUS);
+ break;
+
+ default:
+ internal_error (_("bad resume kind: %d."), zetp->resume_state);
+ }
+
+ /* When stepping over a breakpoint, we need to suppress the breakpoint
+ exception we would otherwise get immediately.
+
+ This requires breakpoints to be already inserted when this function
+ is called. It also handles permanent breakpoints. */
+ if (is_at_breakpoint (tp))
+ cr0[0] |= (1 << INTELGT_CR0_0_BREAKPOINT_SUPPRESS);
+
+ intelgt_write_cr0 (regcache, 0, cr0[0]);
+ intelgt_write_cr0 (regcache, 1, cr0[1]);
+ intelgt_write_cr0 (regcache, 2, cr0[2]);
+
+ dprintf ("thread %s (%s) resumed, cr0.0=%" PRIx32 " .1=%" PRIx32
+ " .2=%" PRIx32 ".", tp->id.to_string ().c_str (),
+ ze_thread_id_str (zetp->id).c_str (), cr0[0], cr0[1], cr0[2]);
+}
+
+void
+intelgt_ze_target::add_regset (target_desc *tdesc, const ze_device_info &dinfo,
+ const zet_debug_regset_properties_t ®prop,
+ long ®num, ze_regset_info_t ®sets,
+ expedite_t &expedite)
+{
+ tdesc_feature *feature = nullptr;
+ const ze_device_properties_t &device = dinfo.properties;
+
+ ze_regset_info regset = {};
+ regset.type = (uint32_t) regprop.type;
+ regset.size = regprop.byteSize;
+ regset.begin = regnum;
+ regset.is_writeable
+ = ((regprop.generalFlags & ZET_DEBUG_REGSET_FLAG_WRITEABLE) != 0);
+
+ if (regprop.count < 1)
+ {
+ warning (_("Ignoring empty regset %u in %s."), regprop.type,
+ device.name);
+ return;
+ }
+
+ if ((regprop.generalFlags & ZET_DEBUG_REGSET_FLAG_READABLE) == 0)
+ {
+ warning (_("Ignoring non-readable regset %u in %s."), regprop.type,
+ device.name);
+ return;
+ }
+
+ switch (regprop.type)
+ {
+ case ZET_DEBUG_REGSET_TYPE_GRF_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_GRF);
+
+ expedite.push_back ("r0");
+
+ intelgt_add_regset (feature, regnum, "r", regprop.count, "grf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_ADDR_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_ADDR);
+
+ intelgt_add_regset (feature, regnum, "a", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 16u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_FLAG_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_FLAG);
+
+ intelgt_add_regset (feature, regnum, "f", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 16u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_CE_INTEL_GPU:
+ /* We expect a single 'ce' register. */
+ if (regprop.count != 1)
+ warning (_("Ignoring %u unexpected 'ce' registers in %s."),
+ regprop.count - 1, device.name);
+
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_CE);
+
+ tdesc_create_reg (feature, "ce", regnum++, 1, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u));
+
+ expedite.push_back ("ce");
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_SR_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_SR);
+
+ expedite.push_back ("sr0");
+ intelgt_add_regset (feature, regnum, "sr", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_CR_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_CR);
+
+ expedite.push_back ("cr0");
+ intelgt_add_regset (feature, regnum, "cr", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_TDR_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_TDR);
+
+ intelgt_add_regset (feature, regnum, "tdr", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 16u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_ACC_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_ACC);
+
+ intelgt_add_regset (feature, regnum, "acc", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_MME_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_MME);
+
+ intelgt_add_regset (feature, regnum, "mme", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_SP_INTEL_GPU:
+ /* We expect a single 'sp' register. */
+ if (regprop.count != 1)
+ warning (_("Ignoring %u unexpected 'sp' registers in %s."),
+ regprop.count - 1, device.name);
+
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_SP);
+
+ tdesc_create_reg (feature, "sp", regnum++, 1, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ regprop.bitSize));
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_SBA_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_SBA);
+
+ switch (regprop.version)
+ {
+ case 0:
+ {
+ const char *regtype = intelgt_uint_reg_type (feature,
+ regprop.bitSize,
+ regprop.bitSize);
+ const char *sbaregs[] = {
+ "genstbase",
+ "sustbase",
+ "dynbase",
+ "iobase",
+ "isabase",
+ "blsustbase",
+ "blsastbase",
+ "btbase",
+ "scrbase",
+ "scrbase2",
+ nullptr
+ };
+ int reg = 0;
+ for (; (reg < regprop.count) && (sbaregs[reg] != nullptr); ++reg)
+ {
+ if ((strcmp (sbaregs[reg], "genstbase") == 0)
+ || (strcmp (sbaregs[reg], "isabase") == 0))
+ {
+ expedite.push_back (sbaregs[reg]);
+ }
+
+ tdesc_create_reg (feature, sbaregs[reg], regnum++, 1,
+ "virtual", regprop.bitSize, regtype);
+ }
+ }
+ break;
+
+ default:
+ warning (_("Ignoring unknown SBA regset version %u in %s"),
+ regprop.version, device.name);
+ break;
+ }
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_DBG_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_DBG);
+
+ intelgt_add_regset (feature, regnum, "dbg", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_FC_INTEL_GPU:
+ feature = tdesc_create_feature (tdesc, intelgt::FEATURE_FC);
+
+ intelgt_add_regset (feature, regnum, "fc", regprop.count, "arf",
+ regprop.bitSize,
+ intelgt_uint_reg_type (feature, regprop.bitSize,
+ 32u),
+ expedite);
+ break;
+
+ case ZET_DEBUG_REGSET_TYPE_INVALID_INTEL_GPU:
+ case ZET_DEBUG_REGSET_TYPE_FORCE_UINT32:
+ break;
+ }
+
+ if (feature == nullptr)
+ {
+ warning (_("Ignoring unknown regset %u in %s."), regprop.type,
+ device.name);
+
+ return;
+ }
+
+ regset.end = regnum;
+ regsets.push_back (regset);
+}
+
+
+/* The Intel GT target ops object. */
+
+static intelgt_ze_target the_intelgt_ze_target;
+
+extern void initialize_low ();
+void
+initialize_low ()
+{
+ /* Delayed initialization of Level-Zero targets. See ze-low.h. */
+ the_intelgt_ze_target.init ();
+ set_target_ops (&the_intelgt_ze_target);
+}
diff --git a/gdbserver/ze-low.cc b/gdbserver/ze-low.cc
new file mode 100644
index 0000000000000000000000000000000000000000..d3dfb1f175ba8b406a9f9d927c71ca6ea5154d2e
--- /dev/null
+++ b/gdbserver/ze-low.cc
@@ -0,0 +1,2996 @@
+/* Target interface for Level-Zero based targets for gdbserver.
+ See https://github.com/oneapi-src/level-zero.git.
+
+ Copyright (C) 2020-2025 Free Software Foundation, Inc.
+
+ This file is part of GDB.
+
+ 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 "ze-low.h"
+#include "dll.h"
+
+#include <level_zero/zet_api.h>
+#include <exception>
+#include <sstream>
+#include <iomanip>
+#include <cstring> /* For snprintf. */
+#include <thread>
+#include <utility>
+#include <algorithm>
+#include <set>
+
+#ifndef USE_WIN32API
+# include <signal.h>
+# include <fcntl.h>
+#endif
+
+
+/* Convenience macros. */
+
+#define dprintf(...) \
+ do \
+ { \
+ if (debug_threads) \
+ { \
+ debug_printf ("%s: ", __FUNCTION__); \
+ debug_printf (__VA_ARGS__); \
+ debug_printf ("\n"); \
+ } \
+ } \
+ while (0)
+
+#ifndef USE_WIN32API
+/* Async interaction stuff.
+
+ The read/write ends of the pipe registered as waitable file in the
+ event loop. */
+static int ze_event_pipe[2] = { -1, -1 };
+#endif
+
+/* Return whether we're in async mode. */
+
+static bool
+ze_is_async ()
+{
+#ifndef USE_WIN32API
+ return (ze_event_pipe[0] != -1);
+#else
+ return false;
+#endif
+}
+
+/* Get rid of any pending event in the pipe. */
+
+static void
+ze_async_flush ()
+{
+ if (!ze_is_async ())
+ return;
+
+#ifndef USE_WIN32API
+ int ret;
+ char buf;
+
+ errno = 0;
+ do
+ ret = read (ze_event_pipe[0], &buf, 1);
+ while (ret >= 0 || (ret == -1 && errno == EINTR));
+#else
+ error (_("%s: tbd"), __FUNCTION__);
+#endif
+}
+
+/* Put something in the pipe, so the event loop wakes up. */
+
+static void
+ze_async_mark ()
+{
+ if (!ze_is_async ())
+ return;
+
+#ifndef USE_WIN32API
+ int ret;
+
+ ze_async_flush ();
+
+ errno = 0;
+ do
+ ret = write (ze_event_pipe[1], "+", 1);
+ while (ret == 0 || (ret == -1 && errno == EINTR));
+
+ /* Ignore EAGAIN. If the pipe is full, the event loop will already
+ be awakened anyway. */
+#else
+ error (_("%s: tbd"), __FUNCTION__);
+#endif
+}
+
+/* Return a human-readable device thread id component string. */
+
+static std::string
+ze_thread_id_component_str (uint32_t component)
+{
+ if (component == UINT32_MAX)
+ return std::string ("all");
+
+ return std::to_string (component);
+}
+
+/* See ze-low.h. */
+
+std::string
+ze_thread_id_str (const ze_device_thread_t &thread)
+{
+ std::stringstream sstream;
+ sstream << ze_thread_id_component_str (thread.slice)
+ << "."
+ << ze_thread_id_component_str (thread.subslice)
+ << "."
+ << ze_thread_id_component_str (thread.eu)
+ << "."
+ << ze_thread_id_component_str (thread.thread);
+
+ return sstream.str ();
+}
+
+/* Return a human-readable UUID string. */
+
+static std::string
+uuid_str (const uint8_t uuid[], size_t size)
+{
+ std::stringstream sstream;
+ while (size--)
+ sstream << std::setw (2) << uuid[size];
+
+ return sstream.str ();
+}
+
+/* Return a human-readable device UUID string. */
+
+static std::string
+driver_uuid_str (const ze_driver_uuid_t &uuid)
+{
+ return uuid_str (uuid.id, sizeof (uuid.id));
+}
+
+/* Return a human-readable process state string. */
+
+static const char *
+ze_process_state_str (ze_process_state state)
+{
+ switch (state)
+ {
+ case ZE_PROCESS_VISIBLE:
+ return "visible";
+
+ case ZE_PROCESS_HIDDEN:
+ return "hidden";
+ }
+
+ return "unknown";
+}
+
+/* Return the pid for DEVICE. */
+
+static int
+ze_device_pid (const ze_device_info &device)
+{
+ if (device.process != nullptr)
+ return device.process->pid;
+
+ return 0;
+}
+
+/* Return the device for PROCESS. */
+
+static ze_device_info *
+ze_process_device (process_info *process)
+{
+ if (process == nullptr)
+ return nullptr;
+
+ process_info_private *zeproc = process->priv;
+ if (zeproc == nullptr)
+ return nullptr;
+
+ return zeproc->device;
+}
+
+/* Return the device for THREAD. */
+
+ze_device_info *
+ze_thread_device (const thread_info *thread)
+{
+ if (thread == nullptr)
+ return nullptr;
+
+ return ze_process_device (thread->process ());
+}
+
+/* Returns whether ID is in SET. */
+
+static bool
+ze_device_thread_in (ze_device_thread_t id, ze_device_thread_t set)
+{
+ if ((set.slice != UINT32_MAX) && (set.slice != id.slice))
+ return false;
+
+ if ((set.subslice != UINT32_MAX) && (set.subslice != id.subslice))
+ return false;
+
+ if ((set.eu != UINT32_MAX) && (set.eu != id.eu))
+ return false;
+
+ if ((set.thread != UINT32_MAX) && (set.thread != id.thread))
+ return false;
+
+ return true;
+}
+
+/* Call FUNC for each thread on DEVICE matching ID. */
+
+template <typename Func>
+static void
+for_each_thread (const ze_device_info &device, ze_device_thread_t id,
+ Func func)
+{
+ device.process->for_each_thread ([id, func] (thread_info *tp)
+ {
+ ze_device_thread_t tid = ze_thread_id (tp);
+ if (ze_device_thread_in (tid, id))
+ func (tp);
+ });
+}
+
+/* Add a process for DEVICE. */
+
+static process_info *
+ze_add_process (ze_device_info *device, ze_process_state state)
+{
+ gdb_assert (device != nullptr);
+
+ process_info *process = add_process (device->ordinal, 1);
+ process->priv = new process_info_private (device, state);
+ process->tdesc = device->tdesc.get ();
+ device->process = process;
+
+ /* Enumerate threads on the device we attached to.
+
+ We debug the entire device so we can enumerate all threads at once. They
+ will be idle some of the time and we won't be able to interact with them.
+ When work gets submitted to the device, the thread dispatcher will
+ distribute the work onto device threads.
+
+ The alternative of only representing threads that are currently executing
+ work would be too intrusive as we'd need to stop each thread on every
+ dispatch. */
+ long tid = 0;
+ uint32_t slice, sslice, eu, thread;
+ const ze_device_properties_t &properties = device->properties;
+ for (slice = 0; slice < properties.numSlices; ++slice)
+ for (sslice = 0; sslice < properties.numSubslicesPerSlice; ++sslice)
+ for (eu = 0; eu < properties.numEUsPerSubslice; ++eu)
+ for (thread = 0; thread < properties.numThreadsPerEU; ++thread)
+ {
+ /* We use the device ordinal as process id. */
+ ptid_t ptid = ptid_t ((int) device->ordinal, ++tid, 0l);
+
+ /* We can only support that many threads. */
+ if (tid < 0)
+ error (_("Too many threads on device %lu: %s."),
+ device->ordinal, properties.name);
+
+ /* Storing the 128b device thread id in the private data. We might
+ want to extend ptid_t and put it there so GDB can show it to the
+ user. */
+ ze_thread_info *zetp = new ze_thread_info {};
+ zetp->id.slice = slice;
+ zetp->id.subslice = sslice;
+ zetp->id.eu = eu;
+ zetp->id.thread = thread;
+
+ /* Assume threads are running until we hear otherwise. */
+ zetp->exec_state = ZE_THREAD_STATE_RUNNING;
+
+ process->add_thread (ptid, zetp);
+ }
+
+ device->nthreads = tid;
+ device->nresumed = tid;
+
+ dprintf ("process %d (%s) with %ld threads created for device %lu: %s.",
+ (int) device->ordinal, ze_process_state_str (state), tid,
+ device->ordinal, properties.name);
+
+ return process;
+}
+
+/* Remove a Level-Zero PROCESS. */
+
+static void
+ze_remove_process (process_info *process)
+{
+ gdb_assert (process != nullptr);
+
+ process->for_each_thread ([=] (thread_info *thread)
+ {
+ delete (ze_thread_info *) thread->target_data ();
+ process->remove_thread (thread);
+ });
+
+ process_info_private *zeinfo = process->priv;
+ gdb_assert (zeinfo != nullptr);
+
+ /* We may or may not have a device.
+
+ When we got detached, we will remove the device first, and remove the
+ process when we select an event from one of its threads.
+
+ When we get a process exit event, the device will remain after the process
+ has been removed. */
+ ze_device_info *device = zeinfo->device;
+ if (device != nullptr)
+ {
+ gdb_assert (device->process == process);
+ device->process = nullptr;
+ }
+
+ process->priv = nullptr;
+ delete zeinfo;
+
+ remove_process (process);
+}
+
+/* Show PROCESS. */
+
+static void
+ze_show_process (process_info *process)
+{
+ gdb_assert (process != nullptr);
+ process_info_private *priv = process->priv;
+
+ gdb_assert (priv != nullptr);
+ switch (priv->state)
+ {
+ case ZE_PROCESS_VISIBLE:
+ return;
+
+ case ZE_PROCESS_HIDDEN:
+ /* FIXME: report state change
+
+ Set priv->status and report the new visibility to GDB. */
+ priv->state = ZE_PROCESS_VISIBLE;
+ return;
+ }
+
+ internal_error (_("unknown process state"));
+}
+
+/* Hide PROCESS. */
+
+static void
+ze_hide_process (process_info *process)
+{
+ gdb_assert (process != nullptr);
+ process_info_private *priv = process->priv;
+
+ gdb_assert (priv != nullptr);
+ switch (priv->state)
+ {
+ case ZE_PROCESS_HIDDEN:
+ return;
+
+ case ZE_PROCESS_VISIBLE:
+ /* FIXME: report state change
+
+ Set priv->status and report the new visibility to GDB. */
+ priv->state = ZE_PROCESS_HIDDEN;
+ return;
+ }
+
+ internal_error (_("unknown process state"));
+}
+
+/* Attach to DEVICE and create a hidden process for it.
+
+ Modifies DEVICE as a side-effect.
+ Returns the created process or nullptr if DEVICE does not support debug. */
+
+static process_info *
+ze_attach (ze_device_info *device)
+{
+ gdb_assert (device != nullptr);
+
+ if (device->session != nullptr)
+ error (_("Already attached to %s."), device->properties.name);
+
+ device->debug_attach_state = zetDebugAttach (device->handle, &device->config,
+ &device->session);
+ switch (device->debug_attach_state)
+ {
+ case ZE_RESULT_SUCCESS:
+ if (device->session == nullptr)
+ error (_("Bad handle returned by zetDebugAttach on %s."),
+ device->properties.name);
+
+ return ze_add_process (device, ZE_PROCESS_HIDDEN);
+
+ case ZE_RESULT_NOT_READY:
+ /* We're too early. The Level-Zero user-mode driver has not been
+ initialized, yet. */
+ error (_("Attempting to attach too early to %s."),
+ device->properties.name);
+
+ case ZE_RESULT_ERROR_UNSUPPORTED_FEATURE:
+ /* Not all sub-devices support attaching to them. */
+ dprintf ("Attach not supported on %s", device->properties.name);
+ return nullptr;
+
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ /* Someone else is already attached. This could be us if we already
+ attached to some other sub-device in this device tree. */
+ error (_("Someone is already attached to %s."),
+ device->properties.name);
+
+ default:
+ error (_("Failed to attach to %s (%x)."), device->properties.name,
+ device->debug_attach_state);
+ }
+}
+
+/* Detach from DEVICE. */
+
+static void
+ze_detach (ze_device_info *device)
+{
+ gdb_assert (device != nullptr);
+
+ zet_debug_session_handle_t session = device->session;
+ if (session == nullptr)
+ error (_("Already detached from %s."), device->properties.name);
+
+ dprintf ("device %lu=%s", device->ordinal, device->properties.name);
+
+ ze_result_t status = zetDebugDetach (session);
+ switch (status)
+ {
+ case ZE_RESULT_ERROR_DEVICE_LOST:
+ case ZE_RESULT_SUCCESS:
+ device->session = nullptr;
+ break;
+
+ default:
+ error (_("Failed to detach from %s (%x)."), device->properties.name,
+ status);
+ }
+}
+
+/* Return a human-readable detach reason string. */
+
+static std::string
+ze_detach_reason_str (zet_debug_detach_reason_t reason)
+{
+ switch (reason)
+ {
+ case ZET_DEBUG_DETACH_REASON_INVALID:
+ return std::string (_("invalid"));
+
+ case ZET_DEBUG_DETACH_REASON_HOST_EXIT:
+ return std::string (_("the host process exited"));
+ }
+
+ return std::string (_("unknown"));
+}
+
+/* Return a human-readable module debug information format string. */
+
+static std::string
+ze_debug_info_format_str (zet_module_debug_info_format_t format)
+{
+ switch (format)
+ {
+ case ZET_MODULE_DEBUG_INFO_FORMAT_ELF_DWARF:
+ return std::string (_("DWARF"));
+ }
+
+ return std::string (_("unknown"));
+}
+
+/* Return a human-readable event string. */
+
+static std::string
+ze_event_str (const zet_debug_event_t &event)
+{
+ std::stringstream sstream;
+
+ switch (event.type)
+ {
+ case ZET_DEBUG_EVENT_TYPE_INVALID:
+ sstream << "invalid";
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_DETACHED:
+ sstream << "detached, reason="
+ << ze_detach_reason_str (event.info.detached.reason);
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_PROCESS_ENTRY:
+ sstream << "process entry";
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_PROCESS_EXIT:
+ sstream << "process exit";
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_MODULE_LOAD:
+ sstream << "module load, format="
+ << ze_debug_info_format_str (event.info.module.format)
+ << ", module=[" << std::hex << event.info.module.moduleBegin
+ << "; " << std::hex << event.info.module.moduleEnd
+ << "), addr=" << std::hex << event.info.module.load
+ << ", need-ack: "
+ << ((event.flags & ZET_DEBUG_EVENT_FLAG_NEED_ACK) != 0);
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_MODULE_UNLOAD:
+ sstream << "module unload, format="
+ << ze_debug_info_format_str (event.info.module.format)
+ << ", module=[" << std::hex << event.info.module.moduleBegin
+ << "; " << std::hex << event.info.module.moduleEnd
+ << "), addr=" << std::hex << event.info.module.load;
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_THREAD_STOPPED:
+ sstream << "thread stopped, thread="
+ << ze_thread_id_str (event.info.thread.thread);
+ return sstream.str ();
+
+ case ZET_DEBUG_EVENT_TYPE_THREAD_UNAVAILABLE:
+ sstream << "thread unavailable, thread="
+ << ze_thread_id_str (event.info.thread.thread);
+ return sstream.str ();
+ }
+
+ sstream << "unknown, code=" << event.type;
+ return sstream.str ();
+}
+
+/* Acknowledge an event, if necessary. */
+
+static void
+ze_ack_event (const ze_device_info &device, const zet_debug_event_t &event)
+{
+ /* There is nothing to do for events that do not need acknowledging. */
+ if ((event.flags & ZET_DEBUG_EVENT_FLAG_NEED_ACK) == 0)
+ return;
+
+ ze_result_t status = zetDebugAcknowledgeEvent (device.session, &event);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ break;
+
+ default:
+ error (_("error acknowledging event: %s: %x."),
+ ze_event_str (event).c_str (), status);
+ }
+}
+
+/* Clear TP's resume state. */
+
+static void
+ze_clear_resume_state (thread_info *tp)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ zetp->resume_state = ZE_THREAD_RESUME_NONE;
+}
+
+/* Set TP's resume state from RKIND. */
+
+static void
+ze_set_resume_state (thread_info *tp, resume_kind rkind)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ switch (rkind)
+ {
+ case resume_continue:
+ zetp->resume_state = ZE_THREAD_RESUME_RUN;
+ return;
+
+ case resume_step:
+ zetp->resume_state = ZE_THREAD_RESUME_STEP;
+ return;
+
+ case resume_stop:
+ zetp->resume_state = ZE_THREAD_RESUME_STOP;
+ return;
+ }
+
+ internal_error (_("bad resume kind: %d."), rkind);
+}
+
+/* Return TP's resume state. */
+
+static enum ze_thread_resume_state_t
+ze_resume_state (const thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ if (zetp == nullptr)
+ return ZE_THREAD_RESUME_NONE;
+
+ return zetp->resume_state;
+}
+
+/* Return TP's execution state. */
+
+static enum ze_thread_exec_state_t
+ze_exec_state (const thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ if (zetp == nullptr)
+ return ZE_THREAD_STATE_UNKNOWN;
+
+ return zetp->exec_state;
+}
+
+/* Return whether TP has a stop execution state. */
+
+static bool
+ze_thread_stopped (const thread_info *tp)
+{
+ ze_thread_exec_state_t state = ze_exec_state (tp);
+
+ return ((state == ZE_THREAD_STATE_STOPPED)
+ || (state == ZE_THREAD_STATE_HELD)
+ || (state == ZE_THREAD_STATE_PAUSED));
+}
+
+/* Return whether TP has a pending event. */
+
+static bool
+ze_has_waitstatus (const thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ if (zetp == nullptr)
+ return false;
+
+ return (zetp->waitstatus.kind () != TARGET_WAITKIND_IGNORE);
+}
+
+/* Return whether TP has a pending priority event. */
+
+static bool
+ze_has_priority_waitstatus (const thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ if (zetp == nullptr)
+ return false;
+
+ switch (zetp->waitstatus.kind ())
+ {
+ case TARGET_WAITKIND_IGNORE:
+ case TARGET_WAITKIND_UNAVAILABLE:
+ return false;
+
+ case TARGET_WAITKIND_STOPPED:
+ /* If this thread was stopped via an interrupt, it is an interesting
+ case if GDB wanted it stopped with a stop resume request. */
+ if ((zetp->stop_reason == TARGET_STOPPED_BY_NO_REASON)
+ && (zetp->waitstatus.sig () == GDB_SIGNAL_TRAP))
+ return (zetp->resume_state == ZE_THREAD_RESUME_STOP);
+
+ /* If this thread stopped spuriously, it is not interesting. */
+ if ((zetp->stop_reason == TARGET_STOPPED_BY_NO_REASON)
+ && (zetp->waitstatus.sig () == GDB_SIGNAL_0))
+ return false;
+
+ return true;
+
+ default:
+ return true;
+ }
+}
+
+/* Return TP's waitstatus and clear it in TP. */
+
+static target_waitstatus
+ze_move_waitstatus (thread_info *tp)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ if (zetp == nullptr)
+ return {};
+
+ target_waitstatus status = zetp->waitstatus;
+ zetp->waitstatus.set_ignore ();
+
+ return status;
+}
+
+/* Indicate that we have been detached from the device corresponding to
+ PROCESS. */
+
+static void
+ze_device_detached (process_info *process, zet_debug_detach_reason_t reason)
+{
+ gdb_assert (process != nullptr);
+
+ /* We model getting detached from a device as the corresponding device process
+ exiting with the detach reason as exit status.
+
+ In the first step, we mark all threads of that process exited. We already
+ use the process exit wait status as all threads will exit together.
+
+ In the second step, when one of those threads gets selected for reporting
+ its event, we will remove the process as part of the reporting flow. */
+
+ process->for_each_thread ([reason] (thread_info *tp)
+ {
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ zetp->waitstatus.set_exited ((int) reason);
+ });
+}
+
+/* Find the regset containing REGNO on DEVICE or throw if not found. */
+
+static ze_regset_info
+ze_find_regset (const ze_device_info &device, long regno)
+{
+ for (const ze_regset_info ®set : device.regsets)
+ {
+ if (regno < regset.begin)
+ continue;
+
+ if (regset.end <= regno)
+ continue;
+
+ return regset;
+ }
+
+ error (_("No register %ld on %s."), regno, device.properties.name);
+}
+
+/* Fetch all registers for THREAD on DEVICE into REGCACHE. */
+
+static void
+ze_fetch_all_registers (const ze_device_info &device,
+ const ze_device_thread_t thread,
+ regcache *regcache)
+{
+ for (const ze_regset_info ®set : device.regsets)
+ {
+ gdb_assert (regset.begin <= regset.end);
+ long lnregs = regset.end - regset.begin;
+
+ gdb_assert (lnregs < UINT32_MAX);
+ uint32_t nregs = (uint32_t) lnregs;
+
+ std::vector<uint8_t> buffer (regset.size * nregs);
+ ze_result_t status
+ = zetDebugReadRegisters (device.session, thread, regset.type, 0,
+ nregs, buffer.data ());
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ {
+ size_t offset = 0;
+ long reg = regset.begin;
+
+ for (; reg < regset.end; reg += 1, offset += regset.size)
+ {
+ if (status == ZE_RESULT_SUCCESS)
+ supply_register (regcache, reg, &buffer[offset]);
+ else
+ supply_register (regcache, reg, nullptr);
+ }
+ }
+ break;
+
+ default:
+ warning (_("Error %x reading regset %" PRIu32 " for %s on %s."),
+ status, regset.type, ze_thread_id_str (thread).c_str (),
+ device.properties.name);
+
+ break;
+ }
+ }
+}
+
+/* Fetch register REGNO for THREAD on DEVICE into REGCACHE. */
+
+static void
+ze_fetch_register (const ze_device_info &device,
+ const ze_device_thread_t thread,
+ regcache *regcache, long regno)
+{
+ ze_regset_info regset = ze_find_regset (device, regno);
+
+ gdb_assert (regset.begin <= regno);
+ long lrsno = regno - regset.begin;
+
+ gdb_assert (lrsno <= UINT32_MAX);
+ uint32_t rsno = (uint32_t) lrsno;
+
+ std::vector<uint8_t> buffer (regset.size);
+ ze_result_t status
+ = zetDebugReadRegisters (device.session, thread, regset.type, rsno, 1,
+ buffer.data ());
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ supply_register (regcache, regno, buffer.data ());
+ break;
+
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ supply_register (regcache, regno, nullptr);
+ break;
+
+ default:
+ warning (_("Error %x reading register %ld (regset %" PRIu32
+ ") for %s on %s."), status, regno, regset.type,
+ ze_thread_id_str (thread).c_str (), device.properties.name);
+ break;
+ }
+}
+
+/* Store all registers for THREAD on DEVICE from REGCACHE. */
+
+static void
+ze_store_all_registers (const ze_device_info &device,
+ const ze_device_thread_t thread,
+ regcache *regcache)
+{
+ for (const ze_regset_info ®set : device.regsets)
+ {
+ if (!regset.is_writeable)
+ continue;
+
+ gdb_assert (regset.begin <= regset.end);
+ long lnregs = regset.end - regset.begin;
+
+ gdb_assert (lnregs < UINT32_MAX);
+ uint32_t nregs = (uint32_t) lnregs;
+
+ std::vector<uint8_t> buffer (regset.size * nregs);
+ size_t offset = 0;
+ long reg = regset.begin;
+ for (; reg < regset.end; reg += 1, offset += regset.size)
+ collect_register (regcache, reg, &buffer[offset]);
+
+ ze_result_t status
+ = zetDebugWriteRegisters (device.session, thread, regset.type, 0,
+ nregs, buffer.data ());
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ break;
+
+ default:
+ error (_("Error %x writing regset %" PRIu32 " for %s on %s."),
+ status, regset.type, ze_thread_id_str (thread).c_str (),
+ device.properties.name);
+ }
+ }
+}
+
+/* Store register REGNO for THREAD on DEVICE from REGCACHE. */
+
+static void
+ze_store_register (const ze_device_info &device,
+ const ze_device_thread_t thread,
+ regcache *regcache, long regno)
+{
+ ze_regset_info regset = ze_find_regset (device, regno);
+
+ if (!regset.is_writeable)
+ error (_("Writing read-only register %ld (regset %" PRIu32
+ ") for %s on %s."), regno, regset.type,
+ ze_thread_id_str (thread).c_str (), device.properties.name);
+
+ gdb_assert (regset.begin <= regno);
+ long lrsno = regno - regset.begin;
+
+ gdb_assert (lrsno <= UINT32_MAX);
+ uint32_t rsno = (uint32_t) lrsno;
+
+ std::vector<uint8_t> buffer (regset.size);
+ collect_register (regcache, regno, buffer.data ());
+
+ ze_result_t status
+ = zetDebugWriteRegisters (device.session, thread, regset.type, rsno, 1,
+ buffer.data ());
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ break;
+
+ default:
+ error (_("Error %x writing register %ld (regset %" PRIu32
+ ") for %s on %s."), status, regno, regset.type,
+ ze_thread_id_str (thread).c_str (),
+ device.properties.name);
+ }
+}
+
+/* Discard TP's regcache. */
+
+static void
+ze_discard_regcache (thread_info *tp)
+{
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ false);
+ gdb_assert (regcache != nullptr);
+
+ regcache->registers_fetched = false;
+ regcache->reset (REG_UNKNOWN);
+}
+
+/* Prepare for resuming TP. Return true if TP should be actually
+ resumed. */
+
+static bool
+ze_prepare_for_resuming (thread_info *tp)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ /* We should not call this function if there is a priority
+ waitstatus. */
+ gdb_assert (!ze_has_priority_waitstatus (tp));
+
+ /* When we get detached, we will remove the device but we will also mark
+ each thread exited. We shouldn't try to resume them. */
+ ze_device_info *device = ze_thread_device (tp);
+ gdb_assert (device != nullptr);
+
+ ze_thread_exec_state_t state = zetp->exec_state;
+ switch (state)
+ {
+ case ZE_THREAD_STATE_PAUSED:
+ zetp->exec_state = ZE_THREAD_STATE_STOPPED;
+
+ [[fallthrough]];
+ case ZE_THREAD_STATE_STOPPED:
+ device->nresumed++;
+ if (device->nresumed > device->nthreads)
+ {
+ device->nresumed = device->nthreads;
+ dprintf ("capping device %lu's nresumed at %ld (all)",
+ device->ordinal, device->nthreads);
+ }
+ return true;
+
+ case ZE_THREAD_STATE_HELD:
+ gdb_assert_not_reached ("threads with 'held' state should "
+ "have been turned into 'stopped'");
+
+ case ZE_THREAD_STATE_UNAVAILABLE:
+ device->nresumed++;
+ if (device->nresumed > device->nthreads)
+ {
+ device->nresumed = device->nthreads;
+ dprintf ("capping device %lu's nresumed at %ld (all)",
+ device->ordinal, device->nthreads);
+ }
+
+ zetp->exec_state = ZE_THREAD_STATE_RUNNING;
+
+ /* Ignore resuming unavailable threads. */
+ return false;
+
+ case ZE_THREAD_STATE_RUNNING:
+ /* Ignore resuming already running threads. */
+ return false;
+
+ case ZE_THREAD_STATE_UNKNOWN:
+ warning (_("thread %s has unknown execution "
+ "state"), tp->id.to_string ().c_str ());
+ return false;
+ }
+
+ internal_error (_("bad execution state: %d."), state);
+}
+
+/* Prepare for stopping TP. Return true if TP should be
+ actually stopped by sending an interrupt to the target. */
+
+static bool
+ze_prepare_for_stopping (thread_info *tp)
+{
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ /* When we get detached, we will remove the device but we will also mark
+ each thread exited. We shouldn't try to stop them. */
+ ze_device_info *device = ze_thread_device (tp);
+ gdb_assert (device != nullptr);
+
+ ze_thread_exec_state_t state = zetp->exec_state;
+ switch (state)
+ {
+ case ZE_THREAD_STATE_STOPPED:
+ /* We silently ignore already stopped threads. */
+ return false;
+
+ case ZE_THREAD_STATE_HELD:
+ gdb_assert_not_reached ("threads with 'held' state should "
+ "have been turned into 'stopped'");
+
+ case ZE_THREAD_STATE_PAUSED:
+ /* A paused thread is already stopped. */
+ zetp->exec_state = ZE_THREAD_STATE_STOPPED;
+ return false;
+
+ case ZE_THREAD_STATE_UNAVAILABLE:
+ case ZE_THREAD_STATE_RUNNING:
+ return true;
+
+ case ZE_THREAD_STATE_UNKNOWN:
+ warning (_("thread %s has unknown execution state"),
+ tp->id.to_string ().c_str ());
+ return false;
+ }
+
+ internal_error (_("bad execution state: %d."), state);
+}
+
+/* Resume THREAD on DEVICE. */
+
+static void
+ze_resume (ze_device_info &device, ze_device_thread_t thread)
+{
+ dprintf ("device %lu=%s, thread=%s", device.ordinal,
+ device.properties.name, ze_thread_id_str (thread).c_str ());
+
+ ze_result_t status = zetDebugResume (device.session, thread);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ break;
+
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ /* Ignore this if we're not modeling DEVICE as a process anymore. */
+ if (device.process == nullptr)
+ break;
+
+ /* The thread is already running or unavailable.
+
+ Assuming our thread state tracking is correct, the thread isn't
+ running, so we assume it became unavailable. That is strange,
+ too, as we had it stopped. */
+ warning (_("thread %s unexpectedly unavailable on %s."),
+ ze_thread_id_str (thread).c_str (), device.properties.name);
+
+ /* Update our thread state to reflect the target. */
+ for_each_thread (device, thread, [&] (thread_info *tp)
+ {
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ zetp->exec_state = ZE_THREAD_STATE_UNAVAILABLE;
+ zetp->waitstatus.set_unavailable ();
+ });
+ break;
+
+ default:
+ error (_("Failed to resume %s on %s: %x."),
+ ze_thread_id_str (thread).c_str (), device.properties.name,
+ status);
+ }
+}
+
+/* Interrupt THREAD on DEVICE. */
+
+static void
+ze_interrupt (ze_device_info &device, ze_device_thread_t thread)
+{
+ dprintf ("device %lu=%s, thread=%s, nresumed=%ld%s",
+ device.ordinal, device.properties.name,
+ ze_thread_id_str (thread).c_str (), device.nresumed,
+ ((device.nresumed == device.nthreads) ? " (all)" : ""));
+
+ ze_result_t status = zetDebugInterrupt (device.session, thread);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ if (ze_is_thread_id_all (thread))
+ device.ninterrupts++;
+
+ break;
+
+ case ZE_RESULT_NOT_READY:
+ /* We already requested THREAD to be stopped. We do not track
+ requests so let's ignore this. */
+ break;
+
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ /* The thread is already stopped or unavailable.
+
+ Assuming that our state tracking works, update non-stopped
+ threads to reflect that. */
+ for_each_thread (device, thread, [&] (thread_info *tp)
+ {
+ if (ze_thread_stopped (tp))
+ return;
+
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ zetp->exec_state = ZE_THREAD_STATE_UNAVAILABLE;
+ zetp->waitstatus.set_unavailable ();
+ });
+ break;
+
+ default:
+ error (_("Failed to interrupt %s on %s: %x."),
+ ze_thread_id_str (thread).c_str (), device.properties.name,
+ status);
+ }
+}
+
+bool
+ze_target::is_range_stepping (thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ if (ze_thread_stopped (tp)
+ && (zetp->resume_state == ZE_THREAD_RESUME_STEP)
+ && (zetp->stop_reason == TARGET_STOPPED_BY_SINGLE_STEP))
+ {
+ regcache *regcache = get_thread_regcache (tp, /* fetch = */ true);
+ CORE_ADDR pc = read_pc (regcache);
+
+ return ((pc >= zetp->step_range_start)
+ && (pc < zetp->step_range_end));
+ }
+
+ return false;
+}
+
+int
+ze_target::attach_to_device (uint32_t pid, ze_device_handle_t device)
+{
+ ze_device_properties_t properties;
+
+ memset (&properties, 0, sizeof (properties));
+ properties.stype = ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES;
+ properties.pNext = nullptr;
+
+ ze_result_t status = zeDeviceGetProperties (device, &properties);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain device properties (%x)."),
+ status);
+ return 0;
+ }
+
+ /* We're a bit paranoid. */
+ properties.name[ZE_MAX_DEVICE_NAME-1] = 0;
+
+ int nattached = 0;
+ uint32_t nsub_devices = 0;
+ status = zeDeviceGetSubDevices (device, &nsub_devices, nullptr);
+ if (status != ZE_RESULT_SUCCESS)
+ warning (_("Failed to get number of sub-devices in %s (%x)."),
+ properties.name, status);
+ else if (nsub_devices > 0)
+ {
+ std::vector<ze_device_handle_t> sub_devices (nsub_devices);
+ status = zeDeviceGetSubDevices (device, &nsub_devices,
+ sub_devices.data ());
+ if (status != ZE_RESULT_SUCCESS)
+ warning (_("Failed to enumerate sub-devices in %s (%x)."),
+ properties.name, status);
+ else
+ for (ze_device_handle_t sub_device : sub_devices)
+ nattached += attach_to_device (pid, sub_device);
+ }
+
+ /* If we attached to a sub-device, we're done. We won't be able to attach to
+ a parent device, anymore. */
+ if (nattached > 0)
+ return nattached;
+
+ /* Allow affecting the normal attach behaviour via environment variables by
+ disallowing attaching to devices or sub-devices. */
+ if (properties.flags & ZE_DEVICE_PROPERTY_FLAG_SUBDEVICE)
+ {
+ const char * const disallow_sub_dev
+ = std::getenv ("ZE_GDB_DO_NOT_ATTACH_TO_SUB_DEVICE");
+ if (disallow_sub_dev != nullptr && *disallow_sub_dev != 0)
+ return nattached;
+ }
+ else
+ {
+ const char * const disallow_dev
+ = std::getenv ("ZE_GDB_DO_NOT_ATTACH_TO_DEVICE");
+ if (disallow_dev != nullptr && *disallow_dev != 0)
+ return nattached;
+ }
+
+ uint32_t nregsets = 0;
+ status = zetDebugGetRegisterSetProperties (device, &nregsets, nullptr);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain number of register sets in %s (%x)."),
+ properties.name, status);
+ return nattached;
+ }
+
+ std::vector<zet_debug_regset_properties_t> regsets (nregsets);
+ status = zetDebugGetRegisterSetProperties (device, &nregsets,
+ regsets.data ());
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain register sets in %s (%x)."),
+ properties.name, status);
+ return nattached;
+ }
+
+ /* Check with the actual target implementation whether it supports this kind
+ of device. */
+ if (!is_device_supported (properties, regsets))
+ {
+ dprintf ("skipping unsupported device %s.", properties.name);
+ return nattached;
+ }
+
+ std::unique_ptr<ze_device_info> dinfo { new ze_device_info };
+ dinfo->config.pid = pid;
+ dinfo->handle = device;
+ dinfo->properties = properties;
+
+ ze_pci_ext_properties_t pci_properties {};
+ status = zeDevicePciGetPropertiesExt (device, &pci_properties);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain PCI properties in %s (%x)."),
+ properties.name, status);
+ pci_properties.address.domain = 0;
+ pci_properties.address.bus = 0;
+ pci_properties.address.device = 0;
+ pci_properties.address.function = 0;
+ }
+
+ target_desc *tdesc = create_tdesc (dinfo.get (), regsets,
+ pci_properties);
+ dinfo->tdesc.reset (tdesc);
+
+ unsigned long ordinal = this->ordinal + 1;
+ if (ordinal == 0)
+ internal_error (_("device ordinal overflow."));
+
+ dinfo->ordinal = ordinal;
+
+ try
+ {
+ process_info *process = ze_attach (dinfo.get ());
+ if (process == nullptr)
+ return nattached;
+ }
+ catch (const gdb_exception_error &except)
+ {
+ warning ("%s", except.what ());
+ }
+
+ /* Add the device even if we were not able to attach to allow attempting to
+ attach to it explicitly later on. */
+ devices.push_back (dinfo.release ());
+ this->ordinal = ordinal;
+
+ nattached += 1;
+ return nattached;
+}
+
+int
+ze_target::attach_to_devices (uint32_t pid)
+{
+ uint32_t ndrivers = 0;
+ ze_result_t status = zeDriverGet (&ndrivers, nullptr);
+ if (status != ZE_RESULT_SUCCESS)
+ error (_("Failed to get number of device drivers (%x)."), status);
+
+ std::vector<ze_driver_handle_t> drivers (ndrivers);
+ status = zeDriverGet (&ndrivers, drivers.data ());
+ if (status != ZE_RESULT_SUCCESS)
+ error (_("Failed to enumerate device drivers (%x)."), status);
+
+ int nattached = 0;
+ for (ze_driver_handle_t driver : drivers)
+ {
+ ze_driver_properties_t properties;
+
+ memset (&properties, 0, sizeof (properties));
+ properties.stype = ZE_STRUCTURE_TYPE_DRIVER_PROPERTIES;
+ properties.pNext = nullptr;
+
+ status = zeDriverGetProperties (driver, &properties);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain driver properties (%x)."),
+ status);
+ continue;
+ }
+
+ ze_api_version_t version;
+ status = zeDriverGetApiVersion (driver, &version);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to obtain API version in %s (%x)."),
+ driver_uuid_str (properties.uuid).c_str (),
+ status);
+ continue;
+ }
+
+ switch (ZE_MAJOR_VERSION (version))
+ {
+ case 1:
+ /* We should be OK with all minor versions. */
+ break;
+
+ default:
+ warning (_("Unsupported API version in %s (%x)."),
+ driver_uuid_str (properties.uuid).c_str (),
+ ZE_MAJOR_VERSION (version));
+ continue;
+ }
+
+ uint32_t ndevices = 0;
+ status = zeDeviceGet (driver, &ndevices, nullptr);
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to get number of devices in %s (%x)."),
+ driver_uuid_str (properties.uuid).c_str (),
+ status);
+ continue;
+ }
+
+ std::vector<ze_device_handle_t> devices (ndevices);
+ status = zeDeviceGet (driver, &ndevices, devices.data ());
+ if (status != ZE_RESULT_SUCCESS)
+ {
+ warning (_("Failed to enumerate devices in %s (%x)."),
+ driver_uuid_str (properties.uuid).c_str (),
+ status);
+ continue;
+ }
+
+ dprintf ("scanning driver %s (%" PRIu32 " devices)",
+ driver_uuid_str (properties.uuid).c_str (), ndevices);
+
+ for (ze_device_handle_t device : devices)
+ nattached += attach_to_device (pid, device);
+ }
+
+ return nattached;
+}
+
+uint64_t
+ze_target::fetch_events (ze_device_info &device)
+{
+ /* There are no events if we're not attached. */
+ if (device.session == nullptr)
+ return 0;
+
+ uint64_t nevents = 0;
+ for (;;)
+ {
+ zet_debug_event_t event = {};
+ ze_result_t status = zetDebugReadEvent (device.session, 0ull, &event);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ nevents += 1;
+ break;
+
+ case ZE_RESULT_NOT_READY:
+ return nevents;
+
+ default:
+ error (_("error fetching events from %s: %x."),
+ device.properties.name, status);
+ }
+
+ dprintf ("received event from device %lu: %s", device.ordinal,
+ ze_event_str (event).c_str ());
+
+ switch (event.type)
+ {
+ case ZET_DEBUG_EVENT_TYPE_INVALID:
+ break;
+
+ case ZET_DEBUG_EVENT_TYPE_DETACHED:
+ {
+ process_info *process = device.process;
+ if (process != nullptr)
+ ze_device_detached (process, event.info.detached.reason);
+
+ /* We're detached, now. */
+ device.session = nullptr;
+ }
+ return nevents;
+
+ case ZET_DEBUG_EVENT_TYPE_PROCESS_ENTRY:
+ ze_ack_event (device, event);
+ ze_show_process (device.process);
+ continue;
+
+ case ZET_DEBUG_EVENT_TYPE_PROCESS_EXIT:
+ ze_ack_event (device, event);
+ ze_hide_process (device.process);
+ continue;
+
+ case ZET_DEBUG_EVENT_TYPE_MODULE_LOAD:
+ {
+ /* We would not remain attached without a process. */
+ process_info *process = device.process;
+ gdb_assert (process != nullptr);
+
+ bool need_ack
+ = ((event.flags & ZET_DEBUG_EVENT_FLAG_NEED_ACK) != 0);
+ loaded_dll (process, event.info.module.moduleBegin,
+ event.info.module.moduleEnd,
+ event.info.module.load, need_ack);
+
+ /* If Level-Zero is not requesting the event to be
+ acknowledged, we're done.
+
+ This happens when attaching to an already running process,
+ for example. We will receive module load events for
+ modules that have already been loaded.
+
+ No need to inform GDB, either, as we expect GDB to query
+ shared libraries after attach. */
+ if (!need_ack)
+ continue;
+
+ device.ack_pending.emplace_back (event);
+
+ /* Loading a new module is a process event. We do not want to
+ overwrite other process events, however, as module loads
+ can also be communicated as part of other events. */
+ process_info_private *zeproc = process->priv;
+ gdb_assert (zeproc != nullptr);
+
+ if (zeproc->waitstatus.kind () != TARGET_WAITKIND_IGNORE)
+ continue;
+
+ /* We use UNAVAILABLE rather than LOADED as the latter implies
+ that the target has stopped. */
+ zeproc->waitstatus.set_unavailable ();
+ }
+ continue;
+
+ case ZET_DEBUG_EVENT_TYPE_MODULE_UNLOAD:
+ {
+ /* We would not remain attached without a process. */
+ process_info *process = device.process;
+ gdb_assert (process != nullptr);
+
+ unloaded_dll (process, event.info.module.moduleBegin,
+ event.info.module.moduleEnd,
+ event.info.module.load);
+
+ /* We don't need an ack, here, but maybe Level-Zero does. */
+ ze_ack_event (device, event);
+
+ /* We do not notify GDB immediately about the module unload.
+ This is harmless until we reclaim the memory for something
+ else. In our case, this can only be another module and we
+ will notify GDB in that case. */
+ }
+ continue;
+
+ case ZET_DEBUG_EVENT_TYPE_THREAD_STOPPED:
+ {
+ ze_device_thread_t tid = event.info.thread.thread;
+ ze_ack_event (device, event);
+
+ /* We would not remain attached without a process. */
+ process_info *process = device.process;
+ gdb_assert (process != nullptr);
+
+ uint32_t nstopped = 0;
+ for_each_thread (device, tid, [&] (thread_info *tp)
+ {
+ /* Ignore threads we know to be stopped.
+
+ We already analyzed the stop reason and probably
+ destroyed it in the process. */
+ if (ze_thread_stopped (tp))
+ return;
+
+ /* Prevent underflowing. */
+ if (device.nresumed > 0)
+ device.nresumed--;
+
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ /* Discard any registers we may have fetched while TP was
+ unavailable. */
+ ze_discard_regcache (tp);
+ try
+ {
+ gdb_signal signal = GDB_SIGNAL_0;
+
+ /* If this is an unavailable thread with a 'stop'
+ resume state, from GDB's point of view the
+ thread was interrupted. In all-stop mode, we
+ keep the event held to not confuse GDB.
+
+ Do the state update before get_stop_reason
+ below, so that in case we access memory, we
+ will do that using the right thread
+ context. */
+ if (!non_stop
+ && (zetp->exec_state == ZE_THREAD_STATE_UNAVAILABLE)
+ && (zetp->resume_state == ZE_THREAD_RESUME_STOP))
+ zetp->exec_state = ZE_THREAD_STATE_HELD;
+ else
+ zetp->exec_state = ZE_THREAD_STATE_STOPPED;
+
+ target_stop_reason reason = get_stop_reason (tp, signal);
+
+ zetp->stop_reason = reason;
+ zetp->waitstatus.set_stopped (signal);
+ nstopped += 1;
+ }
+ /* FIXME: exceptions
+
+ We'd really like to catch some 'thread_unavailable'
+ exception rather than assuming that any exception is
+ due to thread availability. */
+ catch (...)
+ {
+ zetp->exec_state = ZE_THREAD_STATE_UNAVAILABLE;
+ zetp->waitstatus.set_unavailable ();
+ }
+ });
+
+ dprintf ("device %lu's nresumed=%ld%s",
+ device.ordinal, device.nresumed,
+ ((device.nresumed == device.nthreads) ? " (all)" : ""));
+
+ /* This is the response to an interrupt if TID is "all". */
+ if (ze_is_thread_id_all (tid))
+ {
+ if (device.ninterrupts > 0)
+ device.ninterrupts--;
+ else
+ warning (_("ignoring spurious stop-all event on "
+ "device %lu"), device.ordinal);
+ }
+
+ /* A thread event turns a process visible. */
+ if (nstopped > 0)
+ ze_show_process (process);
+ }
+ continue;
+
+ case ZET_DEBUG_EVENT_TYPE_THREAD_UNAVAILABLE:
+ {
+ ze_device_thread_t tid = event.info.thread.thread;
+ ze_ack_event (device, event);
+
+ /* We would not remain attached without a process. */
+ process_info *process = device.process;
+ gdb_assert (process != nullptr);
+
+ for_each_thread (device, tid, [&] (thread_info *tp)
+ {
+ /* Ignore threads we know to be stopped.
+
+ They would not be considered in the response event for
+ an interrupt request. */
+ if (ze_thread_stopped (tp))
+ return;
+
+ /* Prevent underflowing. */
+ if (device.nresumed > 0)
+ device.nresumed--;
+
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ zetp->exec_state = ZE_THREAD_STATE_UNAVAILABLE;
+ zetp->waitstatus.set_unavailable ();
+ });
+
+ dprintf ("device %lu's nresumed=%ld%s",
+ device.ordinal, device.nresumed,
+ ((device.nresumed == device.nthreads) ? " (all)" : ""));
+
+ /* This is the response to an interrupt if TID is "all". */
+ if (ze_is_thread_id_all (tid))
+ {
+ if (device.ninterrupts > 0)
+ device.ninterrupts--;
+ else
+ warning (_("ignoring spurious unavailable-all event on "
+ "device %lu"), device.ordinal);
+ }
+ }
+ continue;
+ }
+
+ /* We only get here if we have not processed EVENT. */
+ warning (_("ignoring event '%s' on %s."),
+ ze_event_str (event).c_str (),
+ device.properties.name);
+
+ /* Acknowledge the ignored event so we don't get stuck. */
+ ze_ack_event (device, event);
+ }
+}
+
+void
+ze_target::fetch_events_all_devices_no_resumed ()
+{
+ uint64_t nresumed = 0;
+ do {
+ nresumed = 0;
+ for (ze_device_info *device : devices)
+ {
+ gdb_assert (device != nullptr);
+
+ /* Ignore devices we're not modelling as processes. */
+ if (device->process == nullptr)
+ continue;
+
+ /* Event processing maintains the number of resumed threads. */
+ fetch_events (*device);
+ nresumed += device->nresumed;
+ }
+ }
+ while (nresumed != 0);
+}
+
+void
+ze_target::init ()
+{
+ ze_result_t status = zeInit (0);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ break;
+
+ default:
+ error (_("Failed to initialize Level-Zero: %x"), status);
+ }
+}
+
+bool
+ze_target::async (bool enable)
+{
+ bool previous = ze_is_async ();
+ if (previous != enable)
+ {
+#ifndef USE_WIN32API
+ if (enable)
+ {
+ try
+ {
+ errno = 0;
+ int status = pipe (ze_event_pipe);
+ if (status == -1)
+ error (_("Failed to create event pipe: %s."),
+ safe_strerror (errno));
+
+ status = fcntl (ze_event_pipe[0], F_SETFL, O_NONBLOCK);
+ if (status == -1)
+ error (_("Failed to set pipe[0] to non-blocking: %s."),
+ safe_strerror (errno));
+
+ status = fcntl (ze_event_pipe[1], F_SETFL, O_NONBLOCK);
+ if (status == -1)
+ error (_("Failed to set pipe[1] to non-blocking: %s."),
+ safe_strerror (errno));
+
+ /* Register the event loop handler. */
+ add_file_handler (ze_event_pipe[0],
+ handle_target_event, NULL,
+ "ze-low");
+
+ /* Always trigger a wait. */
+ ze_async_mark ();
+ }
+ catch (std::exception &ex)
+ {
+ warning ("%s", ex.what ());
+
+ if (ze_event_pipe[0] != -1)
+ {
+ close (ze_event_pipe[0]);
+ ze_event_pipe[0] = -1;
+ }
+
+ if (ze_event_pipe[1] != -1)
+ {
+ close (ze_event_pipe[1]);
+ ze_event_pipe[1] = -1;
+ }
+ }
+ }
+ else
+ {
+ delete_file_handler (ze_event_pipe[0]);
+
+ close (ze_event_pipe[0]);
+ close (ze_event_pipe[1]);
+ ze_event_pipe[0] = -1;
+ ze_event_pipe[1] = -1;
+ }
+#else
+ error (_("%s: tbd"), __FUNCTION__);
+#endif
+ }
+
+ return previous;
+}
+
+int
+ze_target::create_inferior (const char *program,
+ const std::string &args)
+{
+ /* Level-zero does not support creating inferiors. */
+ return -1;
+}
+
+int
+ze_target::attach (int pid)
+{
+ if (!devices.empty ())
+ error (_("Already attached."));
+
+ uint32_t hostpid = (uint32_t) pid;
+ if ((int) hostpid != pid)
+ error (_("Host process id is not supported."));
+
+ int ndevices = attach_to_devices (hostpid);
+ if (ndevices == 0)
+ error (_("No supported devices found."));
+
+ /* Let's check if we were able to attach to at least one device. */
+ int nattached = 0;
+ std::stringstream sstream;
+ sstream << "Failed to attach to any device.";
+ for (ze_device_info *device : devices)
+ {
+ gdb_assert (device != nullptr);
+ switch (device->debug_attach_state)
+ {
+ case ZE_RESULT_SUCCESS:
+ if (device->session == nullptr)
+ {
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "failed to initialize debug session";
+ continue;
+ }
+
+ /* GDB (and higher layers of gdbserver) expects threads stopped on
+ attach in all-stop mode. In non-stop mode, GDB explicitly
+ sends a stop request. */
+ if (!non_stop)
+ {
+ device->process->for_each_thread ([this] (thread_info *tp)
+ {
+ ze_set_resume_state (tp, resume_stop);
+ bool should_stop = ze_prepare_for_stopping (tp);
+ gdb_assert (should_stop);
+ });
+
+ ze_device_thread_t all = ze_thread_id_all ();
+ ze_interrupt (*device, all);
+ }
+
+ nattached += 1;
+ break;
+ case ZE_RESULT_NOT_READY:
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "attempting to attach too early";
+ break;
+ case ZE_RESULT_ERROR_UNSUPPORTED_FEATURE:
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "attaching is not supported";
+ break;
+ case ZE_RESULT_ERROR_INSUFFICIENT_PERMISSIONS:
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "attaching is not permitted";
+ break;
+ case ZE_RESULT_ERROR_NOT_AVAILABLE:
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "a debugger is already attached";
+ break;
+ default:
+ sstream << "\nDevice " << device->ordinal << " ["
+ << device->properties.name << "] : "
+ << "failed to attach with error code '"
+ << std::hex << device->debug_attach_state
+ << std::resetiosflags (std::ios::basefield)
+ << "'";
+ break;
+ }
+ }
+
+ if (nattached == 0)
+ error (_("%s"), sstream.str ().c_str ());
+
+ /* In all-stop mode above, we interrupted the devices. Now we make sure
+ they come to a stop state. So, we fetch events until no device has any
+ resumed threads left. There might be low priority events (e.g.
+ 'module load', 'process entry') we should fetch before fetching higher
+ priority events in the subsequent call of 'wait ()'. If not done here,
+ we fetch the lower priority events in 'wait ()', report an UNAVAILABLE
+ status to GDB and then fetch the higher priority events in 'pause_all'.
+ In a live attach scenario, we don't receive a 'continue' resume request
+ and would miss the thread stopped event. */
+ if (!non_stop)
+ fetch_events_all_devices_no_resumed ();
+
+ /* Return the ID of the last device we attached to. */
+ int device_pid = ze_device_pid (*(devices.back ()));
+ return device_pid;
+}
+
+int
+ze_target::detach (process_info *proc)
+{
+ gdb_assert (proc != nullptr);
+
+ process_info_private *priv = proc->priv;
+ gdb_assert (priv != nullptr);
+
+ ze_device_info *device = priv->device;
+ if (device != nullptr)
+ {
+ /* Resume all the threads on the device. GDB must have already
+ removed all the breakpoints. */
+ try
+ {
+ /* Clear all the pending events first. */
+ proc->for_each_thread ([] (thread_info *tp)
+ {
+ (void) ze_move_waitstatus (tp);
+ });
+
+ resume (*device);
+ }
+ catch (const gdb_exception_error &except)
+ {
+ /* Swallow the error. We are detaching anyway. */
+ dprintf ("%s", except.what ());
+ }
+
+ ze_detach (device);
+ }
+
+ mourn (proc);
+ return 0;
+}
+
+int
+ze_target::kill (process_info *proc)
+{
+ /* Level-zero does not support killing inferiors. */
+ return -1;
+}
+
+void
+ze_target::mourn (process_info *proc)
+{
+ ze_remove_process (proc);
+}
+
+void
+ze_target::join (int pid)
+{
+ /* Nothing to do for Level-Zero targets. */
+}
+
+void
+ze_target::resume (ze_device_info &device)
+{
+ gdb_assert (device.process != nullptr);
+
+ bool has_thread_to_resume = false;
+ device.process->for_each_thread ([&] (thread_info *tp)
+ {
+ ze_set_resume_state (tp, resume_continue);
+ if (ze_prepare_for_resuming (tp))
+ {
+ prepare_thread_resume (tp);
+ regcache_invalidate_thread (tp);
+ has_thread_to_resume = true;
+ }
+ });
+
+ /* There is nothing to resume if nothing is stopped. */
+ if (!has_thread_to_resume)
+ return;
+
+ ze_device_thread_t all = ze_thread_id_all ();
+ ze_resume (device, all);
+}
+
+void
+ze_target::resume_single_thread (thread_info *thread)
+{
+ ze_device_info *device = ze_thread_device (thread);
+ gdb_assert (device != nullptr);
+ ze_thread_info *zetp = ze_thread (thread);
+ gdb_assert (zetp != nullptr);
+
+ bool should_resume = ze_prepare_for_resuming (thread);
+ gdb_assert (should_resume);
+ prepare_thread_resume (thread);
+ regcache_invalidate_thread (thread);
+ ze_resume (*device, zetp->id);
+}
+
+size_t
+ze_target::mark_eventing_threads (ptid_t resume_ptid, resume_kind rkind)
+{
+ /* Note that even if we stopped all, unavailable threads may still
+ report new events as we were not able to stop them.
+
+ We ignore those threads and the unavailable event they report. */
+
+ size_t num_eventing = 0;
+ for_each_thread ([=, &num_eventing] (thread_info *tp)
+ {
+ if (!tp->id.matches (resume_ptid))
+ return;
+
+ if (!ze_has_priority_waitstatus (tp))
+ {
+ (void) ze_move_waitstatus (tp);
+ return;
+ }
+
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ /* If the thread's stop event was being held, it is now the time
+ to convert the state to 'stopped' to unleash the event. */
+ if (zetp->exec_state == ZE_THREAD_STATE_HELD)
+ zetp->exec_state = ZE_THREAD_STATE_STOPPED;
+
+ /* TP may have stopped at a breakpoint that is already deleted
+ by GDB. Consider TP as an eventing thread only if the BP is
+ still there. Because we are inside the 'resume' request, if
+ the BP is valid, GDB must have already re-inserted it.
+
+ FIXME: Keep track of the stop_pc and compare it with the
+ current (i.e. to-be-resumed) pc. */
+ if ((zetp->exec_state == ZE_THREAD_STATE_STOPPED)
+ && (zetp->stop_reason == TARGET_STOPPED_BY_SW_BREAKPOINT)
+ && !is_at_breakpoint (tp))
+ {
+ /* The BP is gone. Clear the waitstatus, too. */
+ target_waitstatus waitstatus = ze_move_waitstatus (tp);
+ if (waitstatus.kind () != TARGET_WAITKIND_STOPPED)
+ warning (_("thread %s has waitstatus %s, expected 'STOPPED'."),
+ tp->id.to_string ().c_str (),
+ waitstatus.to_string ().c_str ());
+ return;
+ }
+
+ /* TP may have stopped during range-stepping, but we reported
+ another thread to GDB. This means the range-stepping state
+ of TP is canceled.
+
+ The condition here is similar to, but not the same as
+ is_range_stepping. We do not check here if the thread's stop
+ pc is within the stepping range. We rather only check if there
+ was a range to step, because the thread may have stopped just
+ when it came out of the range. We should cancel the event in
+ that case, too. */
+ if (ze_thread_stopped (tp)
+ && (zetp->stop_reason == TARGET_STOPPED_BY_SINGLE_STEP)
+ && (zetp->step_range_end > zetp->step_range_start))
+ {
+ target_waitstatus waitstatus = ze_move_waitstatus (tp);
+ dprintf ("Thread %s (%s) was range-stepping, "
+ "canceling the pending event",
+ tp->id.to_string ().c_str (),
+ ze_thread_id_str (zetp->id).c_str ());
+ return;
+ }
+
+ /* Recover the resume state so that the thread can be picked up
+ by 'wait'. */
+ ze_set_resume_state (tp, rkind);
+ num_eventing++;
+ });
+
+ dprintf ("there are %zu eventing threads for ptid %s", num_eventing,
+ resume_ptid.to_string ().c_str ());
+
+ return num_eventing;
+}
+
+/* Display a resume request for logging purposes. */
+
+static void
+print_resume_info (const thread_resume &rinfo)
+{
+ ptid_t rptid = rinfo.thread;
+
+ switch (rinfo.kind)
+ {
+ case resume_continue:
+ dprintf ("received 'continue' resume request for (%s)",
+ rptid.to_string ().c_str ());
+ return;
+
+ case resume_step:
+ dprintf ("received 'step' resume request for (%s)"
+ " in range [0x%" PRIx64 ", 0x%" PRIx64 ")",
+ rptid.to_string ().c_str (),
+ rinfo.step_range_start, rinfo.step_range_end);
+ return;
+
+ case resume_stop:
+ dprintf ("received 'stop' resume request for (%s)",
+ rptid.to_string ().c_str ());
+ return;
+ }
+
+ internal_error (_("bad resume kind: %d."), rinfo.kind);
+}
+
+/* Normalize the resume requests for easier processing later on. */
+
+static void
+normalize_resume_infos (thread_resume *resume_info, size_t n)
+{
+ for (size_t i = 0; i < n; ++i)
+ {
+ thread_resume &rinfo = resume_info[i];
+ ptid_t rptid = rinfo.thread;
+
+ /* Log the original requests. */
+ print_resume_info (rinfo);
+
+ /* We convert ptids of the form (p, -1, 0) to (p, 0, 0) to make
+ 'ptid.matches' work. This transformation is safe because we
+ enumerate the threads starting at 1. */
+ if ((rptid.lwp () == -1) && (rptid.pid () > 0))
+ rinfo.thread = ptid_t (rptid.pid (), 0, 0);
+
+ if (rinfo.sig != 0)
+ {
+ /* Clear out the signal. Our target does not accept
+ signals. */
+ warning (_("Ignoring signal on resume request for %s"),
+ rinfo.thread.to_string ().c_str ());
+ rinfo.sig = 0;
+ }
+ }
+}
+
+/* Resuming threads of a device all at once with a single API call
+ is preferable to resuming threads individually. Therefore, we
+ want to combine individual resume requests with wildcard resumes,
+ if possible.
+
+ For instance, if we receive "vCont;s:1;s:2;c", we would like to
+ make a single ze_resume call with the 'all.all.all.all' thread id
+ after preparing threads 1 and 2 for stepping and the others for
+ continuing.
+
+ We preprocess the resume requests to find for which devices we
+ shall combine the requests. We attempt a merge in all-stop mode
+ when the requests contain continue/step requests only. */
+
+static std::set<ze_device_info *>
+find_wildcard_devices (thread_resume *resume_info, size_t n,
+ const std::list<ze_device_info *> &devices)
+{
+ std::set<ze_device_info *> wildcard_devices;
+
+ if (non_stop)
+ return wildcard_devices;
+
+ for (size_t i = 0; i < n; ++i)
+ {
+ if (resume_info[i].kind == resume_stop)
+ {
+ wildcard_devices.clear ();
+ break;
+ }
+
+ ptid_t rptid = resume_info[i].thread;
+ if (rptid == minus_one_ptid)
+ {
+ for (ze_device_info *device : devices)
+ wildcard_devices.insert (device);
+ break;
+ }
+
+ if (rptid.is_pid ())
+ {
+ process_info *proc = find_process_pid (rptid.pid ());
+ ze_device_info *device = ze_process_device (proc);
+ if (device != nullptr)
+ wildcard_devices.insert (device);
+ }
+ }
+
+ return wildcard_devices;
+}
+
+void
+ze_target::resume (thread_resume *resume_info, size_t n)
+{
+ if (frozen)
+ return;
+
+ /* In all-stop mode, a new resume request overwrites any previous
+ request. We're going to set the request for affected threads below.
+ Clear it for all threads, here.
+
+ In the resume-all case, this will iterate over all threads twice to
+ first clear and then set the resume request. Not ideal, but if we
+ first iterated over all threads to set the resume state, we'd also
+ have to iterate over all threads again in order to actually resume
+ them.
+
+ And if we inverted the loops (i.e. iterate over threads, then over
+ resume requests), we'd miss out on the opportunity to resume all
+ threads at once. */
+ if (!non_stop)
+ for_each_thread ([] (thread_info *tp)
+ {
+ ze_clear_resume_state (tp);
+ });
+
+ normalize_resume_infos (resume_info, n);
+
+ /* Check if there is a thread with a pending event for any of the
+ resume requests. In all-stop mode, we would omit actually
+ resuming the target if there is such a thread. In non-stop mode,
+ we omit resuming the thread itself. */
+ size_t num_eventing = 0;
+ for (size_t i = 0; i < n; ++i)
+ {
+ const thread_resume &rinfo = resume_info[i];
+ resume_kind rkind = rinfo.kind;
+ ptid_t rptid = rinfo.thread;
+
+ if (rkind == resume_stop)
+ continue;
+
+ num_eventing += mark_eventing_threads (rptid, rkind);
+ }
+
+ if ((num_eventing > 0) && !non_stop)
+ return;
+
+ std::set<ze_device_info *> wildcard_devices
+ = find_wildcard_devices (resume_info, n, devices);
+
+ std::set<ze_device_info *> devices_to_resume;
+
+ /* Lambda for applying a resume info on a single thread. */
+ auto apply_resume_info = ([&] (const thread_resume &rinfo,
+ thread_info *tp)
+ {
+ if (ze_has_priority_waitstatus (tp))
+ return;
+
+ ze_set_resume_state (tp, rinfo.kind);
+ ze_device_info *device = ze_thread_device (tp);
+ ze_device_thread_t tid = ze_thread_id (tp);
+
+ switch (rinfo.kind)
+ {
+ case resume_stop:
+ if (ze_prepare_for_stopping (tp))
+ ze_interrupt (*device, tid);
+ break;
+
+ case resume_step:
+ {
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ regcache *regcache
+ = get_thread_regcache (tp, /* fetch = */ true);
+ CORE_ADDR pc = read_pc (regcache);
+
+ /* For single-stepping, start == end. Typically, both are 0.
+ For range-stepping, the PC must be within the range. */
+ CORE_ADDR start = rinfo.step_range_start;
+ CORE_ADDR end = rinfo.step_range_end;
+ gdb_assert ((start == end) || ((pc >= start) && (pc < end)));
+
+ zetp->step_range_start = start;
+ zetp->step_range_end = end;
+ }
+
+ [[fallthrough]];
+ case resume_continue:
+ if (ze_prepare_for_resuming (tp))
+ {
+ prepare_thread_resume (tp);
+ regcache_invalidate_thread (tp);
+
+ /* If the device can be resumed as a whole,
+ omit resuming the thread individually. */
+ if (wildcard_devices.count (device) == 0)
+ ze_resume (*device, tid);
+ else
+ devices_to_resume.insert (device);
+ }
+ break;
+ }
+ });
+
+ /* We may receive multiple requests that apply to a thread. E.g.
+ "vCont;r0xff10,0xffa0:p1.9;c" could be sent to make thread 1.9 do
+ range-stepping from 0xff10 to 0xffa0, while continuing others.
+ According to the Remote Protocol Section E.2 (Packets),
+ "For each inferior thread, the leftmost action with a matching
+ thread-id is applied." For this reason, we keep track of which
+ threads have been resumed individually so that we can skip them
+ when processing wildcard requests.
+
+ Alternatively, we could have the outer loop iterate over threads
+ and the inner loop iterate over resume infos to find the first
+ matching resume info for each thread. There may, however, be a
+ large number of threads and a handful of resume infos that apply
+ to a few threads only. For performance reasons, we prefer to
+ iterate over resume infos in the outer loop. */
+ std::set<thread_info *> individually_resumed_threads;
+ for (size_t i = 0; i < n; ++i)
+ {
+ const thread_resume &rinfo = resume_info[i];
+ gdb_assert (rinfo.sig == 0);
+ ptid_t rptid = rinfo.thread;
+ int rpid = rptid.pid ();
+ if ((rptid == minus_one_ptid)
+ || rptid.is_pid ()
+ || (rptid.lwp () == -1))
+ {
+ for (ze_device_info *device : devices)
+ {
+ gdb_assert (device != nullptr);
+
+ int pid = ze_device_pid (*device);
+ if ((rpid != -1) && (rpid != pid))
+ continue;
+
+ device->process->for_each_thread ([&] (thread_info *tp)
+ {
+ /* We trust that GDB will not send us wildcard resume
+ requests with overlapping pids. Hence, we track
+ only individually-resumed threads. */
+ if (individually_resumed_threads.count (tp) == 0)
+ apply_resume_info (rinfo, tp);
+ });
+ }
+ }
+ else
+ {
+ thread_info *tp = find_thread_ptid (rptid);
+ apply_resume_info (rinfo, tp);
+ individually_resumed_threads.insert (tp);
+ }
+ }
+
+ /* Finally, resume the whole devices. */
+ ze_device_thread_t all = ze_thread_id_all ();
+ for (ze_device_info *device : devices_to_resume)
+ ze_resume (*device, all);
+}
+
+/* Look for a thread preferably with a priority stop
+ event. If we cannot find such an event, we look for an
+ interrupt-related stop event, e.g. a stop because of an
+ external Ctrl-C or an internal pause_all request. We pick
+ a THREAD_UNAVAILABLE event for reporting as the last resort.
+
+ We first make an iteration over the threads to figure out
+ what kind of an event we can report. Once found, we select
+ a thread randomly.
+
+ In all-stop mode, we will ignore unavailable threads when
+ resuming the target. So, unless we explicitly try to interact
+ with them, unavailable threads should be transparent to an
+ all-stop target.
+
+ In non-stop mode, we give more time for unavailable threads to
+ become available and report an event. */
+
+static thread_info *
+ze_find_eventing_thread (ptid_t ptid)
+{
+ using thread_predicate = bool (*) (const thread_info *);
+ thread_predicate is_stopped = [] (const thread_info *tp)
+ {
+ return (ze_thread (tp)->waitstatus.kind () == TARGET_WAITKIND_STOPPED);
+ };
+
+ thread_predicate predicate = nullptr;
+ find_thread (ptid, [&] (thread_info *tp)
+ {
+ /* Only consider threads that were resumed. */
+ ze_thread_resume_state_t state = ze_resume_state (tp);
+ if (state == ZE_THREAD_RESUME_NONE)
+ return false;
+
+ /* If this thread's event is being held, we do not pick it for
+ reporting. */
+ ze_thread_exec_state_t exec_state = ze_exec_state (tp);
+ if (exec_state == ZE_THREAD_STATE_HELD)
+ return false;
+
+ if (ze_has_priority_waitstatus (tp))
+ {
+ predicate = ze_has_priority_waitstatus;
+ return true;
+ }
+
+ if (is_stopped (tp))
+ predicate = is_stopped;
+ else if ((predicate == nullptr) && ze_has_waitstatus (tp))
+ predicate = ze_has_waitstatus;
+
+ return false;
+ });
+
+ thread_info *thread = nullptr;
+ if (predicate != nullptr)
+ thread = find_thread_in_random (ptid, [predicate] (thread_info *tp)
+ {
+ /* Only consider threads that were resumed. */
+ ze_thread_resume_state_t state = ze_resume_state (tp);
+ if (state == ZE_THREAD_RESUME_NONE)
+ return false;
+
+ /* Threads with held events are not picked. */
+ ze_thread_exec_state_t exec_state = ze_exec_state (tp);
+ if (exec_state == ZE_THREAD_STATE_HELD)
+ return false;
+
+ return predicate (tp);
+ });
+ return thread;
+}
+
+ptid_t
+ze_target::wait (ptid_t ptid, target_waitstatus *status,
+ target_wait_flags options)
+{
+ /* We need to wait for further events. */
+ ze_async_mark ();
+
+ do
+ {
+ /* We start by fetching all events.
+
+ This will mark threads stopped and also process solist updates. We may
+ get solist updates even if all device threads are running.
+
+ For all-stop, we anyway want to stop all threads and drain events
+ before reporting the stop to GDB.
+
+ For non-stop, this will allow us to group stop events for multiple
+ threads. */
+ uint64_t nevents;
+ do
+ {
+ nevents = 0;
+
+ for (ze_device_info *device : devices)
+ {
+ gdb_assert (device != nullptr);
+ /* Fetch from any device, regardless of PTID, so that we
+ drain the event queues as much as possible. We use
+ PTID down below to filter the events anyway. */
+ nevents += fetch_events (*device);
+ }
+ }
+ while (nevents > 0);
+
+ /* Next, find a matching entity, whose event we'll report.
+
+ We prioritize process events since they are typically a lot rarer and
+ further have higher impact and should be handled before any thread
+ events of that process.
+
+ Process events are no stop events. They leave threads running,
+ even in all-stop mode. */
+ process_info *process
+ = find_process ([ptid, this] (process_info *proc)
+ {
+ if (!ptid_t (proc->pid).matches (ptid))
+ return false;
+
+ process_info_private *zeproc = proc->priv;
+ gdb_assert (zeproc != nullptr);
+
+ return (zeproc->waitstatus.kind () != TARGET_WAITKIND_IGNORE);
+ });
+
+ /* If we found a process event, it is our primary candidate.
+
+ Process events with a low priority UNAVAILABLE waitstatus do not
+ stop the target in all-stop, but some of its threads might have a
+ pending waitstatus, which requires the stop. If such THREAD is
+ found, we prioritize it, clean the process waitstatus, and fall
+ through to the thread reporting. The process event will
+ piggyback on it.
+
+ We do not take any special care about fairness as we expect process
+ events to be rather rare. */
+ thread_info *thread = nullptr;
+ if (process != nullptr)
+ {
+ process_info_private *zeproc = process->priv;
+ gdb_assert (zeproc != nullptr);
+ ptid_t process_ptid = ptid_t (process->pid);
+
+ /* If we got an unavailable process event, try to find another
+ eventing thread for this process. */
+ if (zeproc->waitstatus.kind () == TARGET_WAITKIND_UNAVAILABLE)
+ thread = ze_find_eventing_thread (process_ptid);
+
+ /* If not found, return the process and clean its waitstatus. */
+ if (thread == nullptr)
+ {
+ *status = zeproc->waitstatus;
+ zeproc->waitstatus.set_ignore ();
+
+ return process_ptid;
+ }
+
+ /* THREAD should always match the PTID: we got a process event,
+ so PTID must be either minus_one or the process's ptid. */
+ gdb_assert (thread->id.matches (ptid));
+
+ /* The process event will piggyback onto the THREAD event.
+ However, we still need to clean the process status. */
+ zeproc->waitstatus.set_ignore ();
+ }
+
+ /* If we have previously found THREAD for the PROCESS, we use it.
+ Otherwise, proceed with searching for a thread event for PTID. */
+ if (thread == nullptr)
+ thread = ze_find_eventing_thread (ptid);
+
+ if (thread != nullptr)
+ {
+ ze_thread_info *zetp = ze_thread (thread);
+ gdb_assert (zetp != nullptr);
+
+ if (is_range_stepping (thread))
+ {
+ /* We are inside the stepping range. Resume the thread
+ and go back to fetching events. */
+ dprintf ("thread %s is stepping in range "
+ "[0x%" PRIx64 ", 0x%" PRIx64 ")",
+ ze_thread_id_str (zetp->id).c_str (),
+ zetp->step_range_start, zetp->step_range_end);
+
+ zetp->waitstatus.set_ignore ();
+ gdb_assert (zetp->resume_state == ZE_THREAD_RESUME_STEP);
+
+ resume_single_thread (thread);
+ continue;
+ }
+
+ /* Resume any thread we didn't want stopped. */
+ if ((zetp->stop_reason == TARGET_STOPPED_BY_NO_REASON)
+ && (zetp->waitstatus.kind () == TARGET_WAITKIND_STOPPED)
+ && (zetp->waitstatus.sig () == GDB_SIGNAL_0))
+ {
+ dprintf ("silently resuming thread %s (%s)",
+ thread->id.to_string ().c_str (),
+ ze_thread_id_str (zetp->id).c_str ());
+
+ /* Undo any previous holding of the event. */
+ zetp->exec_state = ZE_THREAD_STATE_STOPPED;
+ zetp->waitstatus.set_ignore ();
+ ze_set_resume_state (thread, resume_continue);
+
+ resume_single_thread (thread);
+ continue;
+ }
+
+ /* Stop all other threads.
+
+ Save the waitstatus before, because pause_all clears all
+ low-priority events. */
+ *status = zetp->waitstatus;
+
+ if (!non_stop)
+ pause_all (false);
+
+ /* Now also clear the thread's event, regardless of its
+ priority. */
+ zetp->waitstatus.set_ignore ();
+ zetp->step_range_start = 0;
+ zetp->step_range_end = 0;
+
+ /* FIXME: switch_to_thread
+
+ Why isn't the caller switching based on the returned ptid? */
+ switch_to_thread (thread);
+ return thread->id;
+ }
+
+ std::this_thread::yield ();
+ }
+ while ((options & TARGET_WNOHANG) == 0);
+
+ /* We only get here if we did not find any event to report. */
+
+ status->set_ignore ();
+ return null_ptid;
+}
+
+void
+ze_target::fetch_registers (regcache *regcache, int regno)
+{
+ ze_device_thread_t tid = ze_thread_id (regcache->thread);
+ ze_device_info *device = ze_thread_device (regcache->thread);
+ gdb_assert (device != nullptr);
+
+ if (regno == -1)
+ ze_fetch_all_registers (*device, tid, regcache);
+ else
+ ze_fetch_register (*device, tid, regcache, regno);
+}
+
+void
+ze_target::store_registers (regcache *regcache, int regno)
+{
+ ze_device_thread_t tid = ze_thread_id (regcache->thread);
+ ze_device_info *device = ze_thread_device (regcache->thread);
+ gdb_assert (device != nullptr);
+
+ if (regno == -1)
+ ze_store_all_registers (*device, tid, regcache);
+ else
+ ze_store_register (*device, tid, regcache, regno);
+}
+
+/* Determine the thread id and device context for accessing ADDR_SPACE
+ from THREAD. */
+static std::pair<ze_device_thread_t, ze_device_info *>
+ze_memory_access_context (thread_info *thread, unsigned int addr_space)
+{
+ /* With a stopped thread, we can access all address spaces, and we
+ should be able to determine the device for that thread. */
+ if (ze_thread_stopped (thread))
+ return std::pair<ze_device_thread_t, ze_device_info *>
+ { ze_thread_id (thread), ze_thread_device (thread) };
+
+ /* Without a stopped thread, we may only access the default address
+ space and only in the context of thread ALL. */
+ if (addr_space != ZET_DEBUG_MEMORY_SPACE_TYPE_DEFAULT)
+ error (_("need thread to access non-default address space."));
+
+ /* Try to determine the device using THREAD but fall back to the current
+ process' device, e.g. if THREAD is nullptr. */
+ ze_device_info *device = ze_thread_device (thread);
+ if (device == nullptr)
+ {
+ process_info *process = current_process ();
+ device = ze_process_device (process);
+
+ if (device == nullptr)
+ error (_("cannot determine device for memory access."));
+ }
+
+ return std::pair<ze_device_thread_t, ze_device_info *>
+ { ze_thread_id_all (), device };
+}
+
+int
+ze_target::read_memory (thread_info *tp, CORE_ADDR memaddr,
+ unsigned char *myaddr, int len)
+{
+ unsigned int addr_space = 0; /* Only the default space for now. */
+ zet_debug_memory_space_desc_t desc;
+
+ memset (&desc, 0, sizeof (desc));
+ desc.stype = ZET_STRUCTURE_TYPE_DEBUG_MEMORY_SPACE_DESC;
+ desc.pNext = nullptr;
+ desc.type = (zet_debug_memory_space_type_t) addr_space;
+ desc.address = (uint64_t) memaddr;
+
+ std::pair<ze_device_thread_t, ze_device_info *> context
+ = ze_memory_access_context (tp, addr_space);
+ ze_device_thread_t thread = context.first;
+ ze_device_info *device = context.second;
+ gdb_assert (device != nullptr);
+
+ ze_result_t status = zetDebugReadMemory (device->session, thread, &desc,
+ len, myaddr);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ return 0;
+
+ default:
+ dprintf ("error reading %d bytes of memory from %s with %s: %x",
+ len, core_addr_to_string_nz (memaddr),
+ ze_thread_id_str (thread).c_str (), status);
+
+ return EIO;
+ }
+}
+
+int
+ze_target::read_memory (CORE_ADDR memaddr, unsigned char *myaddr, int len)
+{
+ return read_memory (current_thread, memaddr, myaddr, len);
+}
+
+int
+ze_target::write_memory (thread_info *tp, CORE_ADDR memaddr,
+ const unsigned char *myaddr, int len)
+{
+ unsigned int addr_space = 0; /* Only the default space for now. */
+ zet_debug_memory_space_desc_t desc;
+
+ memset (&desc, 0, sizeof (desc));
+ desc.stype = ZET_STRUCTURE_TYPE_DEBUG_MEMORY_SPACE_DESC;
+ desc.pNext = nullptr;
+ desc.type = (zet_debug_memory_space_type_t) addr_space;
+ desc.address = (uint64_t) memaddr;
+
+ std::pair<ze_device_thread_t, ze_device_info *> context
+ = ze_memory_access_context (tp, addr_space);
+ ze_device_thread_t thread = context.first;
+ ze_device_info *device = context.second;
+ gdb_assert (device != nullptr);
+
+ dprintf ("writing %d bytes of memory to %s with %s",
+ len, core_addr_to_string_nz (memaddr),
+ ze_thread_id_str (thread).c_str ());
+
+ ze_result_t status = zetDebugWriteMemory (device->session, thread, &desc,
+ len, myaddr);
+ switch (status)
+ {
+ case ZE_RESULT_SUCCESS:
+ return 0;
+
+ default:
+ dprintf ("error writing %d bytes of memory to %s with %s: %x",
+ len, core_addr_to_string_nz (memaddr),
+ ze_thread_id_str (thread).c_str (), status);
+
+ return EIO;
+ }
+}
+
+int
+ze_target::write_memory (CORE_ADDR memaddr, const unsigned char *myaddr,
+ int len)
+{
+ return write_memory (current_thread, memaddr, myaddr, len);
+}
+
+bool
+ze_target::thread_stopped (struct thread_info *tp)
+{
+ const ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ return (zetp->exec_state == ZE_THREAD_STATE_STOPPED);
+}
+
+void
+ze_target::request_interrupt ()
+{
+ if (current_process () == nullptr)
+ error (_("no current process."));
+
+ process_info *process = current_process ();
+ gdb_assert (process != nullptr);
+
+ process_info_private *priv = process->priv;
+ gdb_assert (priv != nullptr);
+
+ /* The only reason why we would not have a device is if we got detached.
+
+ There is nothing to interrupt in that case. */
+ ze_device_info *device = priv->device;
+ if (device == nullptr)
+ return;
+
+ /* Interrupt is not a resume request. */
+
+ ze_device_thread_t all = ze_thread_id_all ();
+ ze_interrupt (*device, all);
+}
+
+void
+ze_target::pause_all (bool freeze)
+{
+ dprintf ("freeze: %d", freeze);
+
+ if (freeze)
+ {
+ if (frozen == UINT32_MAX)
+ internal_error (_("freeze count overflow"));
+ frozen += 1;
+ }
+
+ /* Nothing to stop if we were frozen already. */
+ if (frozen > 1)
+ return;
+
+ /* Interrupting all threads on devices that have any resumed threads.
+
+ Threads that are already stopped will be ignored by the interrupt. */
+ ze_device_thread_t all = ze_thread_id_all ();
+ for (ze_device_info *device : devices)
+ {
+ gdb_assert (device != nullptr);
+
+ /* Ignore devices we're not modelling as processes. */
+ if (device->process == nullptr)
+ continue;
+
+ if ((device->nresumed != 0) && (device->ninterrupts == 0))
+ ze_interrupt (*device, all);
+ }
+
+ /* Fetch events until no device has any resumed threads left. */
+ fetch_events_all_devices_no_resumed ();
+
+ /* Mark threads we interrupted paused so unpause_all can find then. */
+ for_each_thread ([] (thread_info *tp)
+ {
+ /* A thread without waitstatus has already been processed by a
+ previous pause_all or it has reported its event to higher layers
+ via wait.
+
+ Don't mark it paused. It either already is, if it was stopped by
+ a previous pause_all, or higher layers assume it to be stopped so
+ we don't want it so be resumed by unpause_all. */
+ if (!ze_has_waitstatus (tp))
+ return;
+
+ /* Do not mark threads that wait would pick, even if their event was
+ only just fetched. */
+ if (ze_has_priority_waitstatus (tp))
+ return;
+
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ /* Clear the non-priority waitstatus so wait doesn't pick the thread
+ to report an (unavailable) event we just fetched. */
+ zetp->waitstatus.set_ignore ();
+
+ /* Ignore threads that aren't stopped, most likely because they are
+ unavailable.
+
+ Even though an unavailable thread may have responded to our
+ interrupt, we do not mark it paused because we need to treat
+ unavailable and stopped threads differently in unpause_all. */
+ if (ze_thread_stopped (tp))
+ zetp->exec_state = ZE_THREAD_STATE_PAUSED;
+ });
+}
+
+void
+ze_target::unpause_all (bool unfreeze)
+{
+ dprintf ("freeze: %d", unfreeze);
+
+ if (unfreeze)
+ {
+ if (frozen == 0)
+ internal_error (_("freeze count underflow"));
+ frozen -= 1;
+ }
+
+ /* Nothing to resume if we're still frozen. */
+ if (frozen > 1)
+ return;
+
+ /* Resume threads that were marked by pause_all as well as unavailable
+ threads that were not requested to stop.
+
+ Pause_all leaves the latter marked unavailable. We don't really
+ resume them as they were not actually stopped on the target, but we
+ need to update the thread state and some statistics. */
+
+ /* Check which devices are safe to be resumed and which need to be
+ checked for individual threads to be resumed.
+
+ In all-stop mode, finding a single thread would already block the
+ unpause. We do not expect this to be performance critical (or used
+ at all), however, so let's unify all-stop and non-stop as much as
+ possible. */
+ std::set<ze_device_info *> devices_to_check;
+ std::set<ze_device_info *> devices_to_resume {devices.begin (),
+ devices.end ()};
+
+ for_each_thread ([&] (thread_info *tp)
+ {
+ ze_thread_exec_state_t state = ze_exec_state (tp);
+ switch (state)
+ {
+ case ZE_THREAD_STATE_PAUSED:
+ return;
+
+ case ZE_THREAD_STATE_UNAVAILABLE:
+ {
+ /* Distinguish unavailable threads that we tried to interrupt
+ in pause_all from those that GDB tried to interrupt with a
+ stop resume request. */
+ ze_thread_resume_state_t resume_state = ze_resume_state (tp);
+ if (resume_state != ZE_THREAD_RESUME_STOP)
+ return;
+ }
+
+ [[fallthrough]];
+ case ZE_THREAD_STATE_STOPPED:
+ case ZE_THREAD_STATE_HELD:
+ {
+ ze_device_info *device = ze_thread_device (tp);
+ if (device == nullptr)
+ return;
+
+ devices_to_check.insert (device);
+ devices_to_resume.erase (device);
+ }
+ return;
+
+ case ZE_THREAD_STATE_RUNNING:
+ warning (_("thread %d.%ld running in unpause"), tp->id.pid (),
+ tp->id.lwp ());
+ return;
+
+ case ZE_THREAD_STATE_UNKNOWN:
+ warning (_("thread %d.%ld has unknown execution "
+ "state"), tp->id.pid (), tp->id.lwp ());
+ return;
+ }
+
+ internal_error (_("bad execution state: %d."), state);
+ });
+
+ /* In all-stop mode, any device that cannot be resumed aborts unpause. */
+ if (!non_stop && !devices_to_check.empty ())
+ return;
+
+ /* Resume individual threads.
+
+ In all-stop mode, this will be empty. */
+ for (ze_device_info *device : devices_to_check)
+ {
+ gdb_assert (device != nullptr);
+
+ device->process->for_each_thread ([this] (thread_info *tp)
+ {
+ ze_thread_info *zetp = ze_thread (tp);
+ gdb_assert (zetp != nullptr);
+
+ ze_thread_exec_state_t state = zetp->exec_state;
+ switch (state)
+ {
+ case ZE_THREAD_STATE_STOPPED:
+ case ZE_THREAD_STATE_HELD:
+ case ZE_THREAD_STATE_RUNNING:
+ case ZE_THREAD_STATE_UNKNOWN:
+ /* We already diagnosed unexpected states above. */
+ return;
+
+ case ZE_THREAD_STATE_UNAVAILABLE:
+ {
+ /* Don't touch threads that GDB wants stopped. */
+ ze_thread_resume_state_t resume_state = ze_resume_state (tp);
+ if (resume_state == ZE_THREAD_RESUME_STOP)
+ return;
+
+ /* We don't plan to resume but we still need to prepare TP
+ for nresumed tracking and thread state management. */
+ bool should_resume = ze_prepare_for_resuming (tp);
+ gdb_assert (!should_resume);
+ }
+ return;
+
+ case ZE_THREAD_STATE_PAUSED:
+ resume_single_thread (tp);
+ return;
+ }
+ });
+ }
+
+ /* Resume entire devices at once. */
+ for (ze_device_info *device : devices_to_resume)
+ {
+ gdb_assert (device != nullptr);
+
+ /* Skip devices we're not modeling as processes. */
+ if (device->process == nullptr)
+ continue;
+
+ resume (*device);
+ }
+}
+
+void
+ze_target::ack_library (process_info *process, const char *name)
+{
+ /* All libraries are in-memory. */
+ warning (_("unexpected acknowledgement requested for library %s."), name);
+}
+
+void
+ze_target::ack_in_memory_library (process_info *process,
+ CORE_ADDR begin, CORE_ADDR end)
+{
+ gdb_assert (process != nullptr);
+
+ process_info_private *zeproc = process->priv;
+ gdb_assert (zeproc != nullptr);
+
+ /* The only reason why we would not have a device is if we got detached.
+
+ There is nothing to acknowledge in that case. */
+ ze_device_info *device = zeproc->device;
+ if (device == nullptr)
+ {
+ dprintf ("[%s;%s) device not found.", core_addr_to_string_nz (begin),
+ core_addr_to_string_nz (end));
+ return;
+ }
+
+ events_t &events = device->ack_pending;
+ events_t::iterator it
+ = std::find_if (events.begin (), events.end (),
+ [begin, end] (const zet_debug_event_t &ev)
+ {
+ return ((ev.type == ZET_DEBUG_EVENT_TYPE_MODULE_LOAD)
+ && (ev.info.module.moduleBegin == begin)
+ && (ev.info.module.moduleEnd == end));
+ });
+
+ if (it == events.end ())
+ {
+ dprintf ("[%s;%s) not found.", core_addr_to_string_nz (begin),
+ core_addr_to_string_nz (end));
+ return;
+ }
+
+ ze_ack_event (*device, *it);
+ events.erase (it);
+
+ dprintf ("[%s;%s) acknowledged.", core_addr_to_string_nz (begin),
+ core_addr_to_string_nz (end));
+}
diff --git a/gdbserver/ze-low.h b/gdbserver/ze-low.h
new file mode 100644
index 0000000000000000000000000000000000000000..9f4ce9a1990522e5c09eaa7918708460d0947beb
--- /dev/null
+++ b/gdbserver/ze-low.h
@@ -0,0 +1,496 @@
+/* Target interface for Level-Zero based targets for gdbserver.
+ See https://github.com/oneapi-src/level-zero.git.
+
+ Copyright (C) 2020-2025 Free Software Foundation, Inc.
+
+ This file is part of GDB.
+
+ 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/>. */
+
+#ifndef GDBSERVER_LEVEL_ZERO_LOW_H
+#define GDBSERVER_LEVEL_ZERO_LOW_H
+
+#include "target.h"
+#include "tdesc.h"
+
+#include <level_zero/zet_api.h>
+#include <string>
+#include <vector>
+#include <list>
+
+
+/* Information about register sets reported in target descriptions.
+
+ The main use of this is to find the information relevant for fetching
+ and storing registers via Level-Zero based on register numbers. */
+struct ze_regset_info
+{
+ /* The device-specific Level-Zero register set type. */
+ uint32_t type;
+
+ /* The register size in bytes for reading/writing. */
+ uint32_t size;
+
+ /* The begin (inclusive) and end (exclusive) register numbers for this
+ regset.
+
+ This is used to map register numbers to regset types. */
+ long begin, end;
+
+ /* Whether the regset is writable. We assume all are readable. */
+ bool is_writeable;
+};
+
+/* A vector of regset infos. */
+typedef std::vector<ze_regset_info> ze_regset_info_t;
+
+/* A vector of expedite register names.
+
+ The names are expected to be string literals. The vector must be
+ terminated with a single nullptr entry. */
+typedef std::vector<const char *> expedite_t;
+
+/* A list of debug events. */
+
+typedef std::list<zet_debug_event_t> events_t;
+
+/* Information about devices we're attached to.
+
+ This is pretty similar to process_info. The difference is that we only
+ want to tell GDB about devices that the host application actually uses.
+ To know that, however, we need to attach to all available devices. */
+
+struct ze_device_info
+{
+ /* The debug session configuration. */
+ zet_debug_config_t config = {};
+
+ /* The device handle. This must not be nullptr. */
+ ze_device_handle_t handle = nullptr;
+
+ /* The device's properties. */
+ ze_device_properties_t properties = {};
+
+ /* The debug session handle.
+
+ This is nullptr if we are not currently attached. */
+ zet_debug_session_handle_t session = nullptr;
+
+ /* The state for debug attach attempt.
+
+ This is complementary information for debug session handle. The
+ debug session handle is null, when debug attach attempt fails.
+ In this case, debug attach state contains more information on
+ the last error. */
+ ze_result_t debug_attach_state;
+
+ /* The target description for this device. */
+ target_desc_up tdesc;
+
+ /* The register sets reported in the device's target description. */
+ ze_regset_info_t regsets;
+
+ /* The expedite registers used for this device's target description. */
+ expedite_t expedite;
+
+ /* The device enumeration ordinal number. */
+ unsigned long ordinal = 0;
+
+ /* The process for this device.
+
+ We model devices we're attached to as inferior process. In GDB, we
+ hide inferiors representing devices that are not currently used and
+ only show inferiors for devices that are in use.
+
+ If we are not attached to this device, PROCESS will be nullptr. */
+ process_info *process = nullptr;
+
+ /* A list of to-be-acknowledged events. */
+ events_t ack_pending;
+
+ /* Total number of threads on this device. */
+ unsigned long nthreads = 0;
+
+ /* Number of resumed threads. The value is useful for deciding if
+ we can omit sending an actual interrupt request when we want all
+ threads to be stopped in all-stop mode.
+
+ The value can underflow because of unavailable threads becoming
+ available and generating stop events. Therefore we pay care to
+ prevent underflowing. */
+ unsigned long nresumed = 0;
+
+ /* Number of interrupts sent to this target. */
+ unsigned long ninterrupts = 0;
+};
+
+/* A thread's resume state.
+
+ This is very similar to enum resume_kind except that we need an
+ additional none case to model the thread not being mentioned in any
+ resume request. */
+
+enum ze_thread_resume_state_t
+{
+ /* Gdbserver did not ask anything of this thread. */
+ ZE_THREAD_RESUME_NONE,
+
+ /* The thread shall stop. */
+ ZE_THREAD_RESUME_STOP,
+
+ /* The thread shall run. */
+ ZE_THREAD_RESUME_RUN,
+
+ /* The thread shall step. */
+ ZE_THREAD_RESUME_STEP
+};
+
+/* A thread's execution state. */
+
+enum ze_thread_exec_state_t
+{
+ /* We do not know the thread state. This is likely an error condition. */
+ ZE_THREAD_STATE_UNKNOWN,
+
+ /* The thread is stopped and is expected to remain stopped until we
+ resume it. */
+ ZE_THREAD_STATE_STOPPED,
+
+ /* The thread is stopped but we are holding its stop event until we
+ resume it. */
+ ZE_THREAD_STATE_HELD,
+
+ /* The thread is stopped by pause_all (). In unpause_all (), we need to
+ resume just the paused threads.
+
+ In particular, we need to distinguish threads that reported their
+ event to higher layers in gdbserver and hence have their waitstatus
+ clear (set to ignore) from threads that were paused and had their
+ waitstatus cleared by pause_all ().
+
+ Unavailable threads will not be resumed, so we keep those in state
+ unavailable and only clear their waitstatus to prevent them from
+ getting reported by wait (). */
+ ZE_THREAD_STATE_PAUSED,
+
+ /* The thread is running. We do not know whether it is still available
+ to us and we're able to stop it or whether it would eventually hit a
+ breakpoint.
+
+ When a thread completes executing a kernel it becomes idle and may
+ pick up other workloads, either in this context or in another
+ process' context.
+
+ In the former case, it would still be considered RUNNING from our
+ point of view, even though it started over again with a new set of
+ arguments. In the latter case, it would be UNAVAILABLE. */
+ ZE_THREAD_STATE_RUNNING,
+
+ /* The thread is currently not available to us. It may be idle or it
+ may be executing work on behalf of a different process.
+
+ We cannot distinguish those cases. We're not able to interact with
+ that thread. It may become available again at any time, though.
+
+ From GDB's view, a thread may switch between RUNNING and UNAVAILABLE.
+ We will only know the difference when we try to stop it. It's not
+ entirely clear whether we need to distinguish the two, at all. */
+ ZE_THREAD_STATE_UNAVAILABLE
+};
+
+/* Thread private data for Level-Zero targets. */
+
+struct ze_thread_info
+{
+ /* The thread identifier. */
+ ze_device_thread_t id;
+
+ /* The thread's resume state.
+
+ What does gdbserver want this thread to do. */
+ enum ze_thread_resume_state_t resume_state = ZE_THREAD_RESUME_NONE;
+
+ /* The start/end addresses for range-stepping. */
+ CORE_ADDR step_range_start = 0;
+ CORE_ADDR step_range_end = 0;
+
+ /* The thread's execution state.
+
+ What is this thread actually doing. */
+ enum ze_thread_exec_state_t exec_state = ZE_THREAD_STATE_UNKNOWN;
+
+ /* The thread's stop reason.
+
+ This is only valid if EXEC_STATE == ZE_THREAD_STATE_STOPPED
+ or EXEC_STATE == ZE_THREAD_STATE_HELD. */
+ target_stop_reason stop_reason = TARGET_STOPPED_BY_NO_REASON;
+
+ /* The waitstatus for this thread's last event.
+
+ TARGET_WAITKIND_IGNORE means that there is no last event. */
+ target_waitstatus waitstatus {};
+};
+
+/* Return the ZE thread info for TP. */
+
+static inline
+ze_thread_info *
+ze_thread (thread_info *tp)
+{
+ if (tp == nullptr)
+ return nullptr;
+
+ return (ze_thread_info *) tp->target_data ();
+}
+
+/* Return the ZE thread info for const TP. */
+
+static inline
+const ze_thread_info *
+ze_thread (const thread_info *tp)
+{
+ if (tp == nullptr)
+ return nullptr;
+
+ return (const ze_thread_info *) tp->target_data ();
+}
+
+/* Return the Level-Zero thread id for all threads. */
+
+static inline ze_device_thread_t
+ze_thread_id_all ()
+{
+ ze_device_thread_t all;
+ all.slice = UINT32_MAX;
+ all.subslice = UINT32_MAX;
+ all.eu = UINT32_MAX;
+ all.thread = UINT32_MAX;
+
+ return all;
+}
+
+/* Return true if TID is the all thread id. */
+
+static inline bool
+ze_is_thread_id_all (ze_device_thread_t tid)
+{
+ return (tid.slice == UINT32_MAX
+ && tid.subslice == UINT32_MAX
+ && tid.eu == UINT32_MAX
+ && tid.thread == UINT32_MAX);
+}
+
+/* Return the Level-Zero thread id for THREAD. */
+
+static inline ze_device_thread_t
+ze_thread_id (const thread_info *thread)
+{
+ const ze_thread_info *zetp = ze_thread (thread);
+ if (zetp == nullptr)
+ error (_("No thread."));
+
+ return zetp->id;
+}
+
+/* Return a human-readable device thread id string. */
+
+extern std::string ze_thread_id_str (const ze_device_thread_t &thread);
+
+/* Return the device for THREAD. */
+
+extern ze_device_info *ze_thread_device (const thread_info *thread);
+
+/* The state of a process. */
+
+enum ze_process_state
+{
+ /* The process is visible to the user. */
+ ZE_PROCESS_VISIBLE,
+
+ /* The process is hidden from the user. */
+ ZE_PROCESS_HIDDEN
+};
+
+/* Process info private data for Level-Zero targets. */
+
+struct process_info_private
+{
+ /* The device we're modelling as process.
+
+ In case we get forcefully detached from the device this process
+ represents, DEVICE will be nullptr. The process will remain until
+ the detach event can be reported to GDB. */
+ ze_device_info *device;
+
+ /* The state of this process. */
+ ze_process_state state;
+
+ /* The waitstatus for this process's last event.
+
+ While stop events are reported on threads, module loads and unloads
+ as well as entry and exit are reports on the process itself.
+
+ Neither of these events implies that any of the process' threads
+ stopped or is even available.
+
+ TARGET_WAITKIND_IGNORE means that there is nothing to report. */
+ target_waitstatus waitstatus {};
+
+ process_info_private (ze_device_info *dev, ze_process_state st)
+ : device (dev), state (st)
+ {}
+};
+
+/* Target op definitions for Level-Zero based targets. */
+
+class ze_target : public process_stratum_target
+{
+public:
+ /* Initialize the Level-Zero target.
+
+ We cannot do this inside the ctor since zeInit() would generate a
+ worker thread that would inherit the uninitialized async I/O
+ state.
+
+ Postpone initialization until after async I/O has been
+ initialized. */
+ void init ();
+
+ bool supports_hardware_single_step () override { return true; }
+ bool supports_range_stepping () override { return true; }
+ bool supports_multi_process () override { return true; }
+ bool supports_non_stop () override { return true; }
+ int start_non_stop (bool enable) override { async (enable); return 0; }
+
+ bool async (bool enable) override;
+
+ int create_inferior (const char *program,
+ const std::string &args) override;
+
+ int attach (int pid) override;
+ int detach (process_info *proc) override;
+
+ int kill (process_info *proc) override;
+ void mourn (process_info *proc) override;
+ void join (int pid) override;
+
+ void resume (thread_resume *resume_info, size_t n) override;
+ ptid_t wait (ptid_t ptid, target_waitstatus *status,
+ target_wait_flags options) override;
+
+ void fetch_registers (regcache *regcache, int regno) override;
+ void store_registers (regcache *regcache, int regno) override;
+
+ int read_memory (CORE_ADDR memaddr, unsigned char *myaddr,
+ int len) override;
+
+ int write_memory (CORE_ADDR memaddr, const unsigned char *myaddr,
+ int len) override;
+
+ /* We model h/w threads - they do not exit. */
+ bool thread_alive (ptid_t ptid) override { return true; }
+ bool supports_thread_stopped () override { return true; }
+ bool thread_stopped (struct thread_info *tp) override;
+
+ void request_interrupt () override;
+
+ void pause_all (bool freeze) override;
+ void unpause_all (bool unfreeze) override;
+
+ bool supports_pid_to_exec_file () override { return true; }
+ const char *pid_to_exec_file (int pid) override { return ""; }
+
+ void ack_library (process_info *process, const char *name) override;
+ void ack_in_memory_library (process_info *process, CORE_ADDR begin,
+ CORE_ADDR end) override;
+
+private:
+ typedef std::list<ze_device_info *> devices_t;
+
+ /* The devices we care about. */
+ devices_t devices;
+
+ /* The current device ordinal number used for enumerating devices. */
+ unsigned long ordinal = 0;
+
+ /* The freeze count for pause_all (). */
+ uint32_t frozen = 0;
+
+ /* Attach to PID on devices in the device tree rooted at DEVICE.
+ Returns the number of devices we attached to. */
+ int attach_to_device (uint32_t pid, ze_device_handle_t device);
+
+ /* Attach to all available devices for process PID and store them in
+ this object. Returns the number of devices we attached to. */
+ int attach_to_devices (uint32_t pid);
+
+ /* Fetch and process events from DEVICE. Return number of events. */
+ uint64_t fetch_events (ze_device_info &device);
+
+ /* Fetch events until no device has any resumed threads left. */
+ void fetch_events_all_devices_no_resumed ();
+
+ /* Return the number of threads that match the RESUME_PTID and have
+ new events to report. Also recover these threads' resume state
+ to RKIND. */
+ size_t mark_eventing_threads (ptid_t resume_ptid, enum resume_kind rkind);
+
+ /* Resume all threads on DEVICE. */
+ void resume (ze_device_info &device);
+
+ /* Resume a single thread. This is a helper method that prepares
+ the thread for resuming, invalidates its regcache, and then
+ resumes. The method should be called only when we are sure the
+ thread should be resumed. */
+ void resume_single_thread (thread_info *thread);
+
+ /* Return true if TP has single-stepped within its stepping range. */
+ bool is_range_stepping (thread_info *tp);
+
+protected:
+ /* Check whether a device is supported by this target. */
+ virtual bool is_device_supported
+ (const ze_device_properties_t &,
+ const std::vector<zet_debug_regset_properties_t> &) = 0;
+
+ /* Create a target description for a device and populate the
+ corresponding regset information. */
+ virtual target_desc *create_tdesc
+ (ze_device_info *dinfo,
+ const std::vector<zet_debug_regset_properties_t> &,
+ const ze_pci_ext_properties_t &) = 0;
+
+ /* Return whether TP is at a breakpoint. */
+ virtual bool is_at_breakpoint (thread_info *tp) = 0;
+
+ /* TP stopped. Find out why and return the stop reason. Optionally
+ fill in SIGNAL. */
+ virtual target_stop_reason get_stop_reason (thread_info *tp,
+ gdb_signal &signal) = 0;
+
+ /* Prepare TP for resuming using TP's RESUME_STATE.
+
+ This sets the ze execution state, typically to running. */
+ virtual void prepare_thread_resume (thread_info *tp) = 0;
+
+ /* Read the memory in the context of thread TP. */
+ int read_memory (thread_info *tp, CORE_ADDR memaddr,
+ unsigned char *myaddr, int len);
+
+ /* Write the memory in the context of thread TP. */
+ int write_memory (thread_info *tp, CORE_ADDR memaddr,
+ const unsigned char *myaddr, int len);
+};
+
+#endif /* GDBSERVER_LEVEL_ZERO_LOW_H */
--
2.34.1
Intel Deutschland GmbH
Registered Address: Am Campeon 10, 85579 Neubiberg, Germany
Tel: +49 89 99 8853-0, www.intel.de
Managing Directors: Sean Fennelly, Jeffrey Schneiderman, Tiffany Doon Silva
Chairperson of the Supervisory Board: Nicole Lau
Registered Office: Munich
Commercial Register: Amtsgericht Muenchen HRB 186928
next prev parent reply other threads:[~2025-08-01 9:55 UTC|newest]
Thread overview: 92+ messages / expand[flat|nested] mbox.gz Atom feed top
2025-08-01 9:37 [PATCH v3 00/44] A new target to debug Intel GPUs Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 01/44] gdb, intelgt: add intelgt as a basic machine Tankut Baris Aktemur
2025-12-09 20:44 ` Simon Marchi
2025-12-19 11:13 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 02/44] bfd: add intelgt target to BFD Tankut Baris Aktemur
2025-08-01 12:20 ` Jan Beulich
2025-08-08 5:03 ` Metzger, Markus T
2025-12-09 21:05 ` Simon Marchi
2025-12-19 12:46 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 03/44] ld: add intelgt as a target configuration Tankut Baris Aktemur
2025-08-01 12:06 ` Jan Beulich
2025-08-08 5:03 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 04/44] opcodes: add intelgt as a configuration Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 05/44] gdb, gdbserver, gdbsupport: add 'device' tag to XML target description Tankut Baris Aktemur
2025-12-09 21:27 ` Simon Marchi
2025-12-15 21:03 ` Simon Marchi
2025-12-18 15:04 ` Aktemur, Tankut Baris
2026-01-09 19:12 ` Aktemur, Tankut Baris
2026-01-09 19:34 ` Simon Marchi
2025-08-01 9:37 ` [PATCH v3 06/44] gdb, arch, intelgt: add intelgt arch definitions Tankut Baris Aktemur
2025-12-09 21:48 ` Simon Marchi
2025-12-16 15:47 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 07/44] gdb, intelgt: add the target-dependent definitions for the Intel GT architecture Tankut Baris Aktemur
2025-12-11 18:53 ` Simon Marchi
2025-12-19 16:01 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 08/44] gdb, intelgt: add disassemble feature " Tankut Baris Aktemur
2025-12-11 19:37 ` Simon Marchi
2025-12-23 11:03 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 09/44] gdb, gdbserver, ze: in-memory libraries Tankut Baris Aktemur
2025-12-12 4:13 ` Simon Marchi
2025-12-12 11:20 ` Metzger, Markus T
2025-12-12 19:34 ` Simon Marchi
2025-12-15 13:07 ` Metzger, Markus T
2025-12-15 21:25 ` Simon Marchi
2025-08-01 9:37 ` [PATCH v3 10/44] gdb, gdbserver, rsp, ze: acknowledge libraries Tankut Baris Aktemur
2025-12-12 4:41 ` Simon Marchi
2025-12-12 14:28 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 11/44] gdb, solib, ze: update target_solib_ops::bfd_open_from_target_memory Tankut Baris Aktemur
2025-12-12 4:43 ` Simon Marchi
2025-12-12 14:33 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 12/44] gdb, infrun, ze: allow saving process events Tankut Baris Aktemur
2025-12-12 4:57 ` Simon Marchi
2025-12-15 13:13 ` Metzger, Markus T
2025-12-16 21:10 ` Simon Marchi
2025-12-17 9:30 ` Metzger, Markus T
2025-12-17 20:44 ` Simon Marchi
2025-12-18 7:20 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 13/44] gdb, ze: add TARGET_WAITKIND_UNAVAILABLE Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 14/44] gdb, infrun, ze: handle stopping unavailable threads Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 15/44] gdb, infrun, ze: allow resuming " Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 16/44] gdb, gdbserver, ze: add U stop reply Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 17/44] gdb, gdbserver, ze: add library notification to " Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 18/44] gdbserver, ze: report TARGET_WAITKIND_UNAVAILABLE events Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 19/44] gdb, ze: handle TARGET_WAITKIND_UNAVAILABLE in stop_all_threads Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 20/44] gdb, remote: handle thread unavailability in print_one_stopped_thread Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 21/44] gdb, remote: do 'remote_add_inferior' in 'remote_notice_new_inferior' earlier Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 22/44] gdb, remote: handle a generic process PID in remote_notice_new_inferior Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 23/44] gdb, remote: handle a generic process PID in process_stop_reply Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 24/44] gdb: use the pid from inferior in setup_inferior Tankut Baris Aktemur
2025-12-12 19:51 ` Simon Marchi
2025-12-13 12:40 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 25/44] gdb: revise the pid_to_exec_file target op Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 26/44] gdb: load solibs if the target does not have the notion of an exec file Tankut Baris Aktemur
2025-12-12 20:30 ` Simon Marchi
2026-01-09 19:10 ` Aktemur, Tankut Baris
2025-08-01 9:37 ` [PATCH v3 27/44] gdbserver: import AC_LIB_HAVE_LINKFLAGS macro into the autoconf script Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 28/44] gdbserver: add a pointer to the owner thread in regcache Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 29/44] gdbserver: wait for stopped threads in queue_stop_reply_callback Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 30/44] gdbserver: adjust pid after the target attaches Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 31/44] gdb: do not create a thread after a process event Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 32/44] gdb, ze: on a whole process stop, mark all threads as not_resumed Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 33/44] gdb, dwarf, ze: add DW_OP_INTEL_regval_bits Tankut Baris Aktemur
2025-08-01 12:02 ` Jan Beulich
2025-08-01 12:31 ` Metzger, Markus T
2025-08-01 12:50 ` Jan Beulich
2025-08-08 5:25 ` Metzger, Markus T
2025-08-01 9:37 ` [PATCH v3 34/44] gdbserver: allow configuring for a heterogeneous target Tankut Baris Aktemur
2025-08-01 9:37 ` Tankut Baris Aktemur [this message]
2025-08-01 9:37 ` [PATCH v3 36/44] testsuite, sycl: add SYCL support Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 37/44] testsuite, sycl: add test for backtracing inside a kernel Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 38/44] testsuite, sycl: add test for 'info locals' and 'info args' Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 39/44] testsuite, sycl: add tests for stepping and accessing data elements Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 40/44] testsuite, sycl: add test for 1-D and 2-D parallel_for kernels Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 41/44] testsuite, sycl: add test for scheduler-locking Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 42/44] testsuite, arch, intelgt: add a disassembly test Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 43/44] testsuite, arch, intelgt: add intelgt-program-bp.exp Tankut Baris Aktemur
2025-08-01 9:37 ` [PATCH v3 44/44] testsuite, sycl: test canceling a stepping flow Tankut Baris Aktemur
2025-09-17 12:43 ` [PATCH v3 00/44] A new target to debug Intel GPUs Aktemur, Tankut Baris
2025-10-14 6:34 ` Aktemur, Tankut Baris
2025-12-08 11:32 ` Aktemur, Tankut Baris
2025-12-09 21:30 ` Simon Marchi
2025-12-19 12:52 ` Aktemur, Tankut Baris
Reply instructions:
You may reply publicly to this message via plain-text email
using any one of the following methods:
* Save the following mbox file, import it into your mail client,
and reply-to-all from there: mbox
Avoid top-posting and favor interleaved quoting:
https://en.wikipedia.org/wiki/Posting_style#Interleaved_style
* Reply using the --to, --cc, and --in-reply-to
switches of git-send-email(1):
git send-email \
--in-reply-to=20250801-upstream-intelgt-mvp-v3-35-59ce0f87075b@intel.com \
--to=tankut.baris.aktemur@intel.com \
--cc=config-patches@gnu.org \
--cc=gdb-patches@sourceware.org \
--cc=markus.t.metzger@intel.com \
/path/to/YOUR_REPLY
https://kernel.org/pub/software/scm/git/docs/git-send-email.html
* If your mail client supports setting the In-Reply-To header
via mailto: links, try the mailto: link
Be sure your reply has a Subject: header at the top and a blank line
before the message body.
This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox