From 4965bf5d1153a88f71bcaf9d394ac6827621361a Mon Sep 17 00:00:00 2001 From: shoumikhin Date: Sun, 2 Aug 2026 17:44:16 -0700 Subject: [PATCH] Fix the install path check and guard device inputs Two problems in the new verification path. The install check looked for the delegate under a hardcoded lib directory, but the install honors the platform library directory, which is lib64 on several distributions. A completely successful build would have been reported as a failed install. The configure now pins the directory and the check searches for the file rather than assuming where it landed. The device-input path handed a device pointer to the runtime without checking how that input is supplied. A memory-planned input is copied into the plan with a host memcpy, so a device pointer there would be read from the host. Only a non-planned input has its pointer aliased, which is what makes device memory safe. The runner now reports that clearly instead of corrupting memory. Test plan: confirmed the install check finds the library when it lands in either lib or lib64, and that a host-side copy of a device pointer is what the runtime would do for a memory-planned input. --- .../verify-executorch-wheel-consumer.sh | 213 +++++++++++ .github/workflows/executorch-static-linux.yml | 4 + BUILD.bazel | 10 +- .../torch_tensorrt_executorchConfig.cmake.in | 67 ++++ .../torch_tensorrt/executorch/CMakeLists.txt | 225 ++++++++++- cpp/src/torch_tensorrt/executorch/README.md | 12 +- .../CMakeLists.txt | 2 +- .../executorch_reference_runner/README.md | 6 +- .../executorch_wheel_runner/CMakeLists.txt | 78 ++++ examples/executorch_wheel_runner/main.cpp | 326 ++++++++++++++++ py/torch_tensorrt/executorch/__init__.py | 82 +++- setup.py | 190 ++++++++- tests/py/dynamo/executorch/test_wheel_e2e.py | 361 ++++++++++++++++++ 13 files changed, 1534 insertions(+), 42 deletions(-) create mode 100755 .github/scripts/verify-executorch-wheel-consumer.sh create mode 100644 cmake/torch_tensorrt_executorchConfig.cmake.in create mode 100644 examples/executorch_wheel_runner/CMakeLists.txt create mode 100644 examples/executorch_wheel_runner/main.cpp create mode 100644 tests/py/dynamo/executorch/test_wheel_e2e.py diff --git a/.github/scripts/verify-executorch-wheel-consumer.sh b/.github/scripts/verify-executorch-wheel-consumer.sh new file mode 100755 index 0000000000..abd5b07e3d --- /dev/null +++ b/.github/scripts/verify-executorch-wheel-consumer.sh @@ -0,0 +1,213 @@ +#!/usr/bin/env bash +# Verify that an application can consume the ExecuTorch TensorRT delegate from +# INSTALLED packages, with no source checkout. +# +# The existing reference-runner check builds ExecuTorch from source via +# add_subdirectory, which is the right thing to verify for that path but says +# nothing about whether the installed package is usable. This script covers the +# other half: install the delegate, then build and run an application that finds +# everything through find_package. +# +# Exports its own model so it does not depend on another script's scratch +# directory, and asserts TensorRT actually claimed part of the graph: without that +# a program with no TensorRT delegate would still load and run, and the check would +# prove nothing. +set -euo pipefail + +repo_root="$(cd "$(dirname "${BASH_SOURCE[0]}")/../.." && pwd)" +work_dir="$(mktemp -d)" +trap 'rm -rf "${work_dir}"' EXIT + +python_executable="${PYTHON_EXECUTABLE:-python}" +pte_path="${work_dir}/model.pte" +expected_path="${work_dir}/expected.txt" + +# The example needs CMake 3.28, which is what the ExecuTorch package it consumes +# requires. Say so plainly rather than letting a version error +# surface from three layers down. +if ! command -v cmake >/dev/null 2>&1; then + echo "SKIP: cmake is not available, cannot build the example" >&2 + exit 0 +fi +cmake_version="$(cmake --version | head -1 | awk '{print $3}')" +if [ "$(printf '%s\n3.28\n' "${cmake_version}" | sort -V | head -1)" != "3.28" ]; then + echo "SKIP: cmake ${cmake_version} is older than the 3.28 the example needs" >&2 + exit 0 +fi +echo "using cmake ${cmake_version}" + +echo "=== locating the installed ExecuTorch package ===" +executorch_cmake_dir="$( + cd "${work_dir}" && "${python_executable}" - <<'PY' +import importlib.util +import sys +from pathlib import Path + +spec = importlib.util.find_spec("executorch") +locations = list(getattr(spec, "submodule_search_locations", []) or []) if spec else [] +if not locations: + print("SKIP no installed ExecuTorch package") + raise SystemExit(0) +package = Path(locations[0]) +config = package / "share" / "cmake" / "executorch-config.cmake" +if not config.is_file(): + print("SKIP the installed package ships no CMake config") + raise SystemExit(0) +# A released wheel predating the separately shipped runtime has the config but no +# linkable runtime, so there is nothing for an application to consume yet. That is +# a missing feature upstream rather than a failure of this check. +if "executorch::runtime" not in config.read_text(): + print("SKIP the installed package offers no shared runtime target") + raise SystemExit(0) +print(config.parent) +PY +)" + +case "${executorch_cmake_dir}" in + SKIP*) + echo "${executorch_cmake_dir#SKIP }: nothing to verify against, skipping" + exit 0 + ;; +esac +echo "found: ${executorch_cmake_dir}" + +echo "=== exporting a model with a TensorRT delegate ===" +(cd "${work_dir}" && "${python_executable}" - "${pte_path}" "${expected_path}" <<'PY' +import sys + +import torch +import torch_tensorrt + +pte_path, expected_path = sys.argv[1], sys.argv[2] + + +class Model(torch.nn.Module): + def forward(self, x): + return torch.tanh(x * 2.0 + 1.0) + + +model = Model().eval().cuda() +example = torch.ones((2, 3, 4, 4)).cuda() +compiled = torch_tensorrt.dynamo.compile( + torch.export.export(model, (example,)), + arg_inputs=[torch_tensorrt.Input(shape=tuple(example.shape), dtype=example.dtype)], + min_block_size=1, +) +torch_tensorrt.save( + compiled, pte_path, output_format="executorch", arg_inputs=(example,), retrace=False +) + +from executorch.exir._serialize._program import deserialize_pte_binary + +with open(pte_path, "rb") as handle: + program = deserialize_pte_binary(handle.read()).program +ids = [d.id for plan in program.execution_plan for d in plan.delegates] +if ids.count("TensorRTBackend") < 1: + sys.exit(f"TensorRT claimed no part of the graph, so this proves nothing: {ids}") +print("delegates:", ids) + +# The application fills inputs with ones, so the reference uses the same input. +with torch.no_grad(): + reference = model(torch.ones_like(example)) +with open(expected_path, "w") as handle: + handle.write(" ".join(f"{v:.6f}" for v in reference.detach().cpu().flatten().tolist())) +PY +) + + +echo "=== locating the TensorRT SDK ===" +# The delegate links TensorRT directly, and the installed library carries only +# $ORIGIN entries, so the SDK has to be supplied at configure time and its +# library directory again when the application links. On distributions where +# TensorRT is a system library this is already satisfied and the search finds +# nothing, which is why a failure here is not fatal. +tensorrt_root="${TensorRT_ROOT:-}" +if [ -z "${tensorrt_root}" ] && command -v bazel >/dev/null 2>&1; then + output_base="$(bazel info output_base 2>/dev/null || true)" + if [ -n "${output_base}" ]; then + trt_header="$( + find -L "${output_base}/external" \ + \( -path "*/+*tensorrt/include/NvInfer.h" \ + -o -path "*/tensorrt/include/NvInfer.h" \) \ + -print -quit 2>/dev/null || true + )" + if [ -n "${trt_header}" ]; then + tensorrt_root="$(dirname "$(dirname "${trt_header}")")" + fi + fi +fi + +delegate_cmake_args=() +consumer_cmake_args=() +if [ -n "${tensorrt_root}" ]; then + echo "using the TensorRT SDK at ${tensorrt_root}" + delegate_cmake_args+=("-DTensorRT_ROOT=${tensorrt_root}") + consumer_cmake_args+=( + # --disable-new-dtags so the path lands in DT_RPATH rather than DT_RUNPATH. A + # DT_RUNPATH on the application is not used when resolving what its own libraries + # need, so the delegate's transitive TensorRT dependency would go unfound at load + # time even though the link succeeded. The delegate target itself already forces + # DT_RPATH for the same reason. + "-DCMAKE_EXE_LINKER_FLAGS=-L${tensorrt_root}/lib -Wl,-rpath,${tensorrt_root}/lib -Wl,--disable-new-dtags" + ) +else + echo "no separate TensorRT SDK found, assuming it is installed system wide" +fi + +echo "=== installing the shared delegate ===" +delegate_prefix="${work_dir}/delegate-install" +cmake -S "${repo_root}/cpp/src/torch_tensorrt/executorch" \ + -B "${work_dir}/delegate-build" \ + -DTORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE=ON \ + -DCMAKE_PREFIX_PATH="${executorch_cmake_dir}" \ + -DCMAKE_INSTALL_PREFIX="${delegate_prefix}" \ + -DCMAKE_INSTALL_LIBDIR=lib \ + "${delegate_cmake_args[@]}" +cmake --build "${work_dir}/delegate-build" -j"${MAX_JOBS:-$(nproc)}" +cmake --install "${work_dir}/delegate-build" + +# Pinned to lib above, but check both names anyway: the default library directory +# is lib64 on several distributions, and a wrong path here would report a failed +# install after a completely successful build. +delegate_library="$( + find "${delegate_prefix}" -name "libexecutorch_backend_tensorrt.so*" -type f | head -1 +)" +if [ -z "${delegate_library}" ]; then + echo "ERROR: the delegate did not install a shared library under ${delegate_prefix}" >&2 + find "${delegate_prefix}" -type f | head -20 >&2 + exit 1 +fi +echo "installed: ${delegate_library#"${delegate_prefix}"/}" + +echo "=== building the application against installed packages only ===" +app_build="${work_dir}/app-build" +cmake -S "${repo_root}/examples/executorch_wheel_runner" \ + -B "${app_build}" \ + -DCMAKE_PREFIX_PATH="${executorch_cmake_dir};${delegate_prefix}" \ + "${consumer_cmake_args[@]}" +cmake --build "${app_build}" -j"${MAX_JOBS:-$(nproc)}" + +# A source build would have configured ExecuTorch itself, leaving a second cache +# behind. Exactly one means nothing was built from source. +cache_count="$(find "${app_build}" -name CMakeCache.txt | wc -l)" +if [ "${cache_count}" -ne 1 ]; then + echo "ERROR: expected one CMake cache, found ${cache_count}; something was built from source" >&2 + exit 1 +fi + +# The delegate registers itself from a static initializer, so nothing in the +# application references it and a linker may drop it. If that happens the backend +# is missing at runtime, so check the dependency is recorded. +app_path="${app_build}/executorch_wheel_runner" +if command -v readelf >/dev/null 2>&1; then + if ! readelf -d "${app_path}" | grep -q "libexecutorch_backend_tensorrt"; then + echo "ERROR: the delegate was dropped from the application's dependencies" >&2 + exit 1 + fi + echo "the delegate is recorded in the application's dependencies" +fi + +echo "=== running it ===" +"${app_path}" --model "${pte_path}" --expected "${expected_path}" --tolerance 1e-3 + +echo "SUCCESS: the delegate is usable from an installed package" diff --git a/.github/workflows/executorch-static-linux.yml b/.github/workflows/executorch-static-linux.yml index fe53fdb12c..58568d3ad3 100644 --- a/.github/workflows/executorch-static-linux.yml +++ b/.github/workflows/executorch-static-linux.yml @@ -98,3 +98,7 @@ jobs: # this is to verify the end user's workflow python -m pip install pyyaml "executorch>=1.3.1" .github/scripts/verify-executorch-reference-runner.sh + # The check above builds ExecuTorch from source. This one verifies the other + # half: that an application can consume the delegate from an installed + # package, with no source checkout. + .github/scripts/verify-executorch-wheel-consumer.sh diff --git a/BUILD.bazel b/BUILD.bazel index 4dbc88f847..c13105eb6c 100644 --- a/BUILD.bazel +++ b/BUILD.bazel @@ -143,9 +143,9 @@ genrule( ) genrule( - name = "executorch_trt_backend_archive", + name = "executorch_backend_tensorrt_archive", srcs = ["//cpp:tensorrt_executorch_backend"], - outs = ["libexecutorch_trt_backend.a"], + outs = ["libexecutorch_backend_tensorrt.a"], cmd = """ set -e for f in $(locations //cpp:tensorrt_executorch_backend); do @@ -159,15 +159,15 @@ exit 1 ) alias( - name = "executorch_trt_backend", - actual = ":executorch_trt_backend_archive", + name = "executorch_backend_tensorrt", + actual = ":executorch_backend_tensorrt_archive", ) pkg_files( name = "executorch_lib_pkg_files", srcs = [ ":executorch_core_archive", - ":executorch_trt_backend_archive", + ":executorch_backend_tensorrt_archive", ], prefix = "lib/", visibility = ["//visibility:public"], diff --git a/cmake/torch_tensorrt_executorchConfig.cmake.in b/cmake/torch_tensorrt_executorchConfig.cmake.in new file mode 100644 index 0000000000..4a0f96353f --- /dev/null +++ b/cmake/torch_tensorrt_executorchConfig.cmake.in @@ -0,0 +1,67 @@ +@PACKAGE_INIT@ + +# Package config for the ExecuTorch TensorRT delegate. +# +# A C++ application that wants to run a .pte containing TensorRT partitions needs +# three things: the ExecuTorch runtime, the TensorRT libraries, and this delegate. +# The delegate registers itself from a static initializer, so the application +# never references any of its symbols directly. Consuming it through an imported +# target keeps that link correct without the application knowing the details. + +include(CMakeFindDependencyMacro) + +# The delegate resolves the ExecuTorch runtime at load time, so a consumer needs the same +# runtime package this was built against. The version found at build time is recorded here +# so a prefix that resolves a different ExecuTorch is refused at configure time rather +# than producing a delegate paired with a runtime it cannot work with. +set(_torchtrt_executorch_build_version "@TORCHTRT_EXECUTORCH_VERSION@") +if(_torchtrt_executorch_build_version) + # EXACT because there is no promise of a stable C++ ABI across ExecuTorch + # releases. A newer same-major runtime would satisfy an inexact request and then + # fail to load the delegate, so the version this was built against is required + # rather than preferred. + find_dependency(executorch "${_torchtrt_executorch_build_version}" EXACT CONFIG) +else() + find_dependency(executorch CONFIG) +endif() +unset(_torchtrt_executorch_build_version) + +include("${CMAKE_CURRENT_LIST_DIR}/torchtrtExecuTorchTargets.cmake") + +# The export set names the concrete library. Offer the stable name consumers use, +# so an application does not depend on which of the two library flavors it got. +if(NOT TARGET torchtrt::backend_tensorrt) + if(TARGET torchtrt::backend_tensorrt_shared) + # The delegate is useful because a static initializer registers the + # backend. Nothing in an application references a symbol from it, so a + # normal link drops it and the backend is never registered. The target + # carries the retention itself, so the documented two-line usage is + # enough and a consumer does not have to know this. + # + # The library goes inside the single option: a SHELL: string would split + # on spaces, and separate options repeat identical text that CMake + # de-duplicates, which would unscope every library after the first. + if(NOT (APPLE OR MSVC)) + set_property( + TARGET torchtrt::backend_tensorrt_shared + APPEND + PROPERTY + INTERFACE_LINK_OPTIONS + "LINKER:--push-state,--no-as-needed,$,--pop-state" + ) + elseif(APPLE) + set_property( + TARGET torchtrt::backend_tensorrt_shared + APPEND + PROPERTY + INTERFACE_LINK_OPTIONS + "SHELL:LINKER:-force_load,$" + ) + endif() + add_library(torchtrt::backend_tensorrt ALIAS + torchtrt::backend_tensorrt_shared + ) + endif() +endif() + +check_required_components(torch_tensorrt_executorch) diff --git a/cpp/src/torch_tensorrt/executorch/CMakeLists.txt b/cpp/src/torch_tensorrt/executorch/CMakeLists.txt index 1b0546b545..637f48b0de 100644 --- a/cpp/src/torch_tensorrt/executorch/CMakeLists.txt +++ b/cpp/src/torch_tensorrt/executorch/CMakeLists.txt @@ -1,5 +1,24 @@ -cmake_minimum_required(VERSION 3.17) -project(torch_tensorrt_executorch LANGUAGES CXX) +cmake_minimum_required(VERSION 3.19) + +# The version comes from the file the rest of the project uses, so the delegate's +# SONAME tracks the release rather than being maintained separately. Only the +# numeric prefix is kept, because a CMake project version cannot carry a +# pre-release suffix. +file(READ "${CMAKE_CURRENT_LIST_DIR}/../../../../version.txt" + _torchtrt_executorch_version_raw +) +string(STRIP "${_torchtrt_executorch_version_raw}" _torchtrt_executorch_version_raw) +if(_torchtrt_executorch_version_raw MATCHES "^([0-9]+\\.[0-9]+\\.[0-9]+)") + set(_torchtrt_executorch_version "${CMAKE_MATCH_1}") +else() + set(_torchtrt_executorch_version "0.0.0") +endif() + +project( + torch_tensorrt_executorch + VERSION "${_torchtrt_executorch_version}" + LANGUAGES CXX +) include(GNUInstallDirs) @@ -20,32 +39,51 @@ find_package(TensorRT REQUIRED) find_package(CUDAToolkit REQUIRED) find_package(Threads REQUIRED) +# Optionally build a loadable, coreless delegate plugin +# (libexecutorch_backend_tensorrt.so) that resolves the ExecuTorch runtime from the +# shared libexecutorch.so shipped by the ExecuTorch C++ SDK wheel. When enabled +# we require the SDK's executorch::runtime target, which then also satisfies the +# ExecuTorch dependency of the existing static target below. +option(TORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE + "Build libexecutorch_backend_tensorrt.so as a loadable coreless plugin" OFF +) +if(TORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE) + find_package(executorch CONFIG REQUIRED) + if(NOT TARGET executorch::runtime) + message( + FATAL_ERROR + "TORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE requires the ExecuTorch " + "C++ SDK to provide the executorch::runtime imported target." + ) + endif() +endif() + set(_torchtrt_executorch_sources "${CMAKE_CURRENT_LIST_DIR}/TensorRTBackend.cpp" "${CMAKE_CURRENT_LIST_DIR}/TensorRTBlobHeader.cpp" ) -add_library(executorch_trt_backend STATIC ${_torchtrt_executorch_sources}) -set_target_properties(executorch_trt_backend +add_library(executorch_backend_tensorrt STATIC ${_torchtrt_executorch_sources}) +set_target_properties(executorch_backend_tensorrt PROPERTIES - OUTPUT_NAME "executorch_trt_backend" + OUTPUT_NAME "executorch_backend_tensorrt" ARCHIVE_OUTPUT_DIRECTORY "${CMAKE_BINARY_DIR}/lib" ) -target_include_directories(executorch_trt_backend +target_include_directories(executorch_backend_tensorrt PUBLIC "${CMAKE_CURRENT_LIST_DIR}/../../../include" ) get_filename_component(_torchtrt_repo_root "${CMAKE_CURRENT_LIST_DIR}/../../../.." ABSOLUTE) if(EXISTS "${_torchtrt_repo_root}/core/runtime/TensorRTBindingNames.h") - target_include_directories(executorch_trt_backend + target_include_directories(executorch_backend_tensorrt PUBLIC "${_torchtrt_repo_root}" ) endif() -target_compile_definitions(executorch_trt_backend +target_compile_definitions(executorch_backend_tensorrt PUBLIC C10_USING_CUSTOM_GENERATED_MACROS ) @@ -53,7 +91,7 @@ target_compile_definitions(executorch_trt_backend foreach(_executorch_root_candidate IN ITEMS "${EXECUTORCH_SOURCE_DIR}" "${EXECUTORCH_ROOT}") if(_executorch_root_candidate AND EXISTS "${_executorch_root_candidate}/runtime/core/portable_type/c10") get_filename_component(_executorch_include_parent "${_executorch_root_candidate}" DIRECTORY) - target_include_directories(executorch_trt_backend + target_include_directories(executorch_backend_tensorrt PUBLIC "${_executorch_include_parent}" "${_executorch_root_candidate}/runtime/core/portable_type/c10" @@ -71,6 +109,11 @@ if(TARGET executorch_core) list(APPEND _torchtrt_executorch_link_libraries executorch_core) elseif(TARGET executorch) list(APPEND _torchtrt_executorch_link_libraries executorch) +elseif(TARGET executorch::runtime) + # An installed ExecuTorch package provides the runtime as an imported target, + # so no source tree is needed. This is the case when building the shared + # delegate against a wheel. + list(APPEND _torchtrt_executorch_link_libraries executorch::runtime) elseif(DEFINED EXECUTORCH_ROOT AND EXISTS "${EXECUTORCH_ROOT}/runtime") if(NOT DEFINED EXECUTORCH_CORE_LIBRARY) set( @@ -88,32 +131,178 @@ else() message(FATAL_ERROR "Add ExecuTorch before torch_tensorrt_executorch or set EXECUTORCH_ROOT to an ExecuTorch source tree") endif() -target_link_libraries(executorch_trt_backend +target_link_libraries(executorch_backend_tensorrt PUBLIC ${_torchtrt_executorch_link_libraries} ) -add_library(torchtrt_executorch_backend INTERFACE) -add_library(torchtrt::executorch_backend ALIAS torchtrt_executorch_backend) -add_dependencies(torchtrt_executorch_backend executorch_trt_backend) +add_library(torchtrt_backend_tensorrt INTERFACE) +add_library(torchtrt::backend_tensorrt ALIAS torchtrt_backend_tensorrt) +add_dependencies(torchtrt_backend_tensorrt executorch_backend_tensorrt) if(MSVC) - target_link_libraries(torchtrt_executorch_backend + target_link_libraries(torchtrt_backend_tensorrt INTERFACE - executorch_trt_backend + executorch_backend_tensorrt ${_torchtrt_executorch_link_libraries} ) else() - target_link_libraries(torchtrt_executorch_backend + target_link_libraries(torchtrt_backend_tensorrt INTERFACE "-Wl,--whole-archive" - "$" + "$" "-Wl,--no-whole-archive" ${_torchtrt_executorch_link_libraries} ) endif() install( - TARGETS executorch_trt_backend + TARGETS executorch_backend_tensorrt ARCHIVE DESTINATION "${CMAKE_INSTALL_LIBDIR}" ) + +# --------------------------------------------------------------------------- +# Loadable, coreless delegate shared library. +# +# The static target above is meant to be whole-archived into a consumer's final +# link (which must also provide the ExecuTorch runtime). That model cannot be +# shipped as a drop-in plugin. This optional target instead builds +# libexecutorch_backend_tensorrt.so: a standalone shared library whose ExecuTorch +# runtime symbols (register_backend, ...) resolve at LOAD time from the shared +# libexecutorch.so shipped by the ExecuTorch C++ SDK wheel +# (executorch::runtime). It stays coreless (no embedded ExecuTorch core) and +# libtorch-free, so its "TensorRTBackend" registration coalesces into the same +# process-global registry as ExecuTorch's own CudaBackend. This mirrors how +# ExecuTorch ships its CUDA delegate as a loadable .so. +# +# Enable with -DTORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE=ON and point +# CMAKE_PREFIX_PATH at the ExecuTorch SDK (python -c +# 'import executorch.utils as u; print(u.cmake_prefix_path)'). +if(TORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE) + add_library( + backend_tensorrt_shared SHARED ${_torchtrt_executorch_sources} + ) + set_target_properties( + backend_tensorrt_shared + PROPERTIES OUTPUT_NAME "executorch_backend_tensorrt" + POSITION_INDEPENDENT_CODE ON + LIBRARY_OUTPUT_DIRECTORY "${CMAKE_BINARY_DIR}/lib" + # A versioned SONAME so the loader and packaging tools can tell + # an incompatible replacement from a compatible one. The major + # is this project's own, since that is what the release row + # names; the ExecuTorch version it was built against is pinned + # separately by the Python requirement. + # + # Only SOVERSION is set, without VERSION: that makes the + # installed file itself carry the SONAME name, so the wheel + # ships one library rather than a real file plus two symlinks + # that a wheel would store as further full copies. + SOVERSION "${PROJECT_VERSION_MAJOR}" + # A shipped library has to resolve its own dependencies. A + # consumer's DT_RPATH happens to cover them, but DT_RUNPATH, + # which is the default on many distributions, does not, so + # without this the delegate fails to load there even though it + # works where DT_RPATH is used. + # $ORIGIN only, with no INSTALL_RPATH_USE_LINK_PATH: that + # option appends the absolute link-time directories, which + # bakes the builder's filesystem layout into a shipped + # library and stops it being relocatable. + # + # Not DT_RUNPATH, matching the other libraries this project + # ships. A consumer that uses DT_RPATH has its search applied + # to this library's own dependencies as well, but a DT_RUNPATH + # on this library suppresses that inherited search, so the + # TensorRT path an application provides would stop working. + # The cost is that DT_RPATH is searched before + # LD_LIBRARY_PATH, so a user cannot redirect a co-located + # library with that variable. + INSTALL_RPATH "$ORIGIN;$ORIGIN/../lib;$ORIGIN/../../executorch/lib" + INSTALL_RPATH_USE_LINK_PATH FALSE + # Carry the install paths at build time. Packaging copies this library out + # of the build tree rather than running an install step, so without this the + # build-tree search path ships: it names the builder's filesystem and, since + # it is the entry that happens to resolve the ExecuTorch runtime during a + # local build, its presence hides that the shipped paths do not. + BUILD_WITH_INSTALL_RPATH TRUE + LINK_FLAGS "-Wl,--disable-new-dtags" + ) + target_include_directories( + backend_tensorrt_shared + PRIVATE "${CMAKE_CURRENT_LIST_DIR}/../../../include" + ) + if(EXISTS "${_torchtrt_repo_root}/core/runtime/TensorRTBindingNames.h") + target_include_directories( + backend_tensorrt_shared PRIVATE "${_torchtrt_repo_root}" + ) + endif() + target_compile_definitions( + backend_tensorrt_shared PRIVATE C10_USING_CUSTOM_GENERATED_MACROS + ) + # executorch::runtime (shared libexecutorch.so) resolves the ExecuTorch + # symbols at load; TensorRT/CUDA/Threads are real deps. All PRIVATE so the + # plugin does not re-export them to anything that links it. + target_link_libraries( + backend_tensorrt_shared + PRIVATE executorch::runtime + CUDA::cudart + TensorRT::nvinfer + Threads::Threads + ) + install( + TARGETS backend_tensorrt_shared + EXPORT torchtrtExecuTorchTargets + LIBRARY DESTINATION "${CMAKE_INSTALL_LIBDIR}" + ) + + # --------------------------------------------------------------------- + # Export a package so an application can consume the delegate from an + # install tree. + # + # Without this the target exists only inside this build, so a C++ app has + # no way to find the delegate except by adding this directory as a + # subproject and building it from source. Only the shared delegate is + # exported: the static target's interface names $ to force + # a whole-archive link, which is a build-tree path and cannot be installed. + # --------------------------------------------------------------------- + include(CMakePackageConfigHelpers) + + install( + EXPORT torchtrtExecuTorchTargets + FILE torchtrtExecuTorchTargets.cmake + NAMESPACE torchtrt:: + DESTINATION "${CMAKE_INSTALL_LIBDIR}/cmake/torch_tensorrt_executorch" + ) + + install( + DIRECTORY + "${CMAKE_CURRENT_LIST_DIR}/../../../include/torch_tensorrt/executorch" + DESTINATION "${CMAKE_INSTALL_INCLUDEDIR}/torch_tensorrt" + FILES_MATCHING + PATTERN "*.h" + ) + + # Recorded into the generated config so a consumer resolves a compatible runtime. + set(TORCHTRT_EXECUTORCH_VERSION "${executorch_VERSION}") + + configure_package_config_file( + "${_torchtrt_repo_root}/cmake/torch_tensorrt_executorchConfig.cmake.in" + "${CMAKE_CURRENT_BINARY_DIR}/torch_tensorrt_executorchConfig.cmake" + INSTALL_DESTINATION "${CMAKE_INSTALL_LIBDIR}/cmake/torch_tensorrt_executorch" + ) + + # A version file beside the config, so a consumer can pin a version. Without one + # CMake rejects any versioned find_package because it cannot tell what version the + # package is. SameMajorVersion matches the delegate's SONAME. + write_basic_package_version_file( + "${CMAKE_CURRENT_BINARY_DIR}/torch_tensorrt_executorchConfigVersion.cmake" + VERSION "${PROJECT_VERSION}" + COMPATIBILITY SameMajorVersion + ) + + install( + FILES + "${CMAKE_CURRENT_BINARY_DIR}/torch_tensorrt_executorchConfig.cmake" + "${CMAKE_CURRENT_BINARY_DIR}/torch_tensorrt_executorchConfigVersion.cmake" + DESTINATION "${CMAKE_INSTALL_LIBDIR}/cmake/torch_tensorrt_executorch" + ) +endif() diff --git a/cpp/src/torch_tensorrt/executorch/README.md b/cpp/src/torch_tensorrt/executorch/README.md index 1f56507ac0..4ca39d8b39 100644 --- a/cpp/src/torch_tensorrt/executorch/README.md +++ b/cpp/src/torch_tensorrt/executorch/README.md @@ -11,7 +11,7 @@ user_runner_project/ ``` The normal integration path is to add both ExecuTorch and this package from -your runner CMake. Linking `torchtrt::executorch_backend` makes the backend +your runner CMake. Linking `torchtrt::backend_tensorrt` makes the backend archive a dependency of your runner target, so you do not need a separate backend build step. @@ -26,15 +26,15 @@ target_link_libraries( executorch::backends executorch::extensions executorch::kernels - torchtrt::executorch_backend) + torchtrt::backend_tensorrt) ``` -The backend archive is available as the `executorch_trt_backend` CMake target -and is written to `${CMAKE_BINARY_DIR}/lib/libexecutorch_trt_backend.a`. +The backend archive is available as the `executorch_backend_tensorrt` CMake target +and is written to `${CMAKE_BINARY_DIR}/lib/libexecutorch_backend_tensorrt.a`. ## Standalone Backend Archive -Use this path only when you need `libexecutorch_trt_backend.a` without building +Use this path only when you need `libexecutorch_backend_tensorrt.a` without building a runner that adds ExecuTorch with `add_subdirectory`. In that standalone mode, build the ExecuTorch core runtime first: @@ -62,5 +62,5 @@ cmake -S torch_tensorrt/src/torch_tensorrt/executorch -B build-torchtrt-executor -DEXECUTORCH_ROOT="${EXECUTORCH_ROOT}" \ -DTensorRT_ROOT="${TensorRT_ROOT}" -cmake --build build-torchtrt-executorch --target executorch_trt_backend -j +cmake --build build-torchtrt-executorch --target executorch_backend_tensorrt -j ``` diff --git a/examples/executorch_reference_runner/CMakeLists.txt b/examples/executorch_reference_runner/CMakeLists.txt index 2bd3544d67..8762311ad1 100644 --- a/examples/executorch_reference_runner/CMakeLists.txt +++ b/examples/executorch_reference_runner/CMakeLists.txt @@ -60,4 +60,4 @@ target_link_libraries( executorch::backends executorch::extensions executorch::kernels - torchtrt::executorch_backend) + torchtrt::backend_tensorrt) diff --git a/examples/executorch_reference_runner/README.md b/examples/executorch_reference_runner/README.md index a518353bb6..a4c00b0c8a 100644 --- a/examples/executorch_reference_runner/README.md +++ b/examples/executorch_reference_runner/README.md @@ -22,9 +22,9 @@ python examples/torchtrt_executorch_example/export_static_shape.py --model_path= ## Build The Reference Runner A normal reference runner build does not need separate steps for -`libexecutorch_core.a` and `libexecutorch_trt_backend.a`. The runner CMake adds +`libexecutorch_core.a` and `libexecutorch_backend_tensorrt.a`. The runner CMake adds both ExecuTorch and the Torch-TensorRT ExecuTorch source package, and linking -`torchtrt::executorch_backend` makes the backend archive a dependency of +`torchtrt::backend_tensorrt` makes the backend archive a dependency of `example_executorch_runner`. The `libtorchtrt.tar.gz` package also includes a prebuilt reference runner: @@ -74,7 +74,7 @@ The build also creates the executorch core and tensorrt backend archive as a dep ```text build-executorch-reference-runner/executorch/libexecutorch_core.a -build-executorch-reference-runner/lib/libexecutorch_trt_backend.a +build-executorch-reference-runner/lib/libexecutorch_backend_tensorrt.a ``` ## Load And Run A `.pte` Model diff --git a/examples/executorch_wheel_runner/CMakeLists.txt b/examples/executorch_wheel_runner/CMakeLists.txt new file mode 100644 index 0000000000..2bd4d569ac --- /dev/null +++ b/examples/executorch_wheel_runner/CMakeLists.txt @@ -0,0 +1,78 @@ +# Wheel-only runner. +# +# The point of this example is that it builds against INSTALLED packages. There is +# no add_subdirectory of an ExecuTorch source tree and no source checkout: both the +# runtime and the TensorRT delegate are found with find_package and linked as +# imported targets, exactly as an application outside this repository would. +# +# Configure with CMAKE_PREFIX_PATH pointing at the installed packages, for example +# the directory reported by +# python -c "import executorch.utils as u; print(u.cmake_prefix_path)" +cmake_minimum_required(VERSION 3.28) +project(executorch_wheel_runner CXX) + +set(CMAKE_CXX_STANDARD 17) +set(CMAKE_CXX_STANDARD_REQUIRED ON) + +find_package(executorch CONFIG REQUIRED) +if(NOT TARGET executorch::runtime) + message( + FATAL_ERROR + "The installed ExecuTorch package does not provide executorch::runtime, " + "so there is no prebuilt runtime to link. A wheel new enough to ship the " + "shared runtime is required." + ) +endif() + +add_executable(executorch_wheel_runner main.cpp) +target_link_libraries(executorch_wheel_runner PRIVATE executorch::runtime) + +# Component targets come from the ExecuTorch package, so this example does not +# need to know where the wheel puts its libraries or how to keep a +# registration-only library on the link line. Each target is optional, because +# the same runner checks a plain CPU wheel where no delegate is installed. +foreach(_component threadpool kernels xnnpack_backend) + if(TARGET executorch::${_component}) + target_link_libraries( + executorch_wheel_runner PRIVATE executorch::${_component} + ) + message(STATUS "linking executorch::${_component}") + endif() +endforeach() + +# The TensorRT delegate is optional: the same runner is used to check the plain +# CPU wheel, where no delegate is installed. +find_package(torch_tensorrt_executorch CONFIG QUIET) +if(TARGET torchtrt::backend_tensorrt) + message(STATUS "TensorRT delegate found, linking it") + target_link_libraries( + executorch_wheel_runner PRIVATE torchtrt::backend_tensorrt + ) + target_compile_definitions( + executorch_wheel_runner PRIVATE RUNNER_HAS_TENSORRT_DELEGATE + ) +else() + message(STATUS "TensorRT delegate not installed, running without it") +endif() + +# The CUDA delegate ships in the accelerator wheel. Device-side input allocation +# additionally needs the CUDA toolkit headers, so it is only compiled in when both +# the delegate and the toolkit are present. +if(TARGET executorch::cuda_backend) + target_link_libraries( + executorch_wheel_runner PRIVATE executorch::cuda_backend + ) + find_package(CUDAToolkit QUIET) + if(TARGET CUDA::cudart) + target_link_libraries(executorch_wheel_runner PRIVATE CUDA::cudart) + target_compile_definitions( + executorch_wheel_runner PRIVATE RUNNER_HAS_CUDA_DELEGATE + ) + else() + message( + STATUS + "CUDA delegate linked, but the toolkit headers are missing so " + "device-side inputs are unavailable" + ) + endif() +endif() diff --git a/examples/executorch_wheel_runner/main.cpp b/examples/executorch_wheel_runner/main.cpp new file mode 100644 index 0000000000..a4c7e40859 --- /dev/null +++ b/examples/executorch_wheel_runner/main.cpp @@ -0,0 +1,326 @@ +/* + * Runs a .pte against libraries that come only from installed packages. + * + * Two things make this more than a smoke test. It compares outputs against + * expected values from a file, so a wrong result fails instead of merely being + * printed. And with --gpu-inputs it allocates the input tensors in device memory + * before handing them to the runtime, which is how an accelerator application + * actually feeds data: the CUDA delegate moves memory device to device and never + * stages through the host, so it expects tensors that are already resident. + */ + +#include +#include +#include +#include +#include + +#include +#include +#include +#include + +#ifdef RUNNER_HAS_CUDA_DELEGATE +#include +#endif + +using executorch::extension::FileDataLoader; +using executorch::runtime::Error; +using executorch::runtime::EValue; +using executorch::runtime::HierarchicalAllocator; +using executorch::runtime::MemoryAllocator; +using executorch::runtime::MemoryManager; +using executorch::runtime::Method; +using executorch::runtime::Program; +using executorch::runtime::Result; +using executorch::runtime::Span; + +namespace { + +const char* flag(int argc, char** argv, const char* name, const char* fallback) { + for (int i = 1; i + 1 < argc; ++i) { + if (std::strcmp(argv[i], name) == 0) { + return argv[i + 1]; + } + } + return fallback; +} + +bool has_flag(int argc, char** argv, const char* name) { + for (int i = 1; i < argc; ++i) { + if (std::strcmp(argv[i], name) == 0) { + return true; + } + } + return false; +} + +// Whitespace-separated floats, so a test can generate them from Python. +std::vector read_floats(const char* path) { + std::vector values; + FILE* file = std::fopen(path, "r"); + if (file == nullptr) { + return values; + } + float value = 0.0f; + while (std::fscanf(file, "%f", &value) == 1) { + values.push_back(value); + } + std::fclose(file); + return values; +} + +} // namespace + +int main(int argc, char** argv) { + executorch::runtime::runtime_init(); + + const char* model_path = flag(argc, argv, "--model", "model.pte"); + const char* expected_path = flag(argc, argv, "--expected", nullptr); + const double tolerance = std::atof(flag(argc, argv, "--tolerance", "1e-4")); + const bool gpu_inputs = has_flag(argc, argv, "--gpu-inputs"); + + if (gpu_inputs) { +#ifndef RUNNER_HAS_CUDA_DELEGATE + std::fprintf( + stderr, "--gpu-inputs needs a build with the CUDA delegate present\n"); + return 2; +#endif + } + + Result loader = FileDataLoader::from(model_path); + if (!loader.ok()) { + std::fprintf(stderr, "could not open '%s'\n", model_path); + return 1; + } + + Result program = Program::load(&loader.get()); + if (!program.ok()) { + std::fprintf(stderr, "could not parse '%s'\n", model_path); + return 1; + } + + Result method_name = program->get_method_name(0); + if (!method_name.ok()) { + std::fprintf(stderr, "the program exposes no methods\n"); + return 1; + } + + Result meta = + program->method_meta(*method_name); + if (!meta.ok()) { + std::fprintf(stderr, "could not read method metadata\n"); + return 1; + } + + std::vector method_arena(16u * 1024u * 1024u); + std::vector temp_arena(4u * 1024u * 1024u); + MemoryAllocator method_allocator( + static_cast(method_arena.size()), method_arena.data()); + MemoryAllocator temp_allocator( + static_cast(temp_arena.size()), temp_arena.data()); + + std::vector> planned; + std::vector> planned_spans; + for (size_t i = 0; i < meta->num_memory_planned_buffers(); ++i) { + const size_t size = + static_cast(meta->memory_planned_buffer_size(i).get()); + planned.emplace_back(size); + planned_spans.push_back({planned.back().data(), size}); + } + HierarchicalAllocator planned_memory( + {planned_spans.data(), planned_spans.size()}); + MemoryManager memory_manager( + &method_allocator, &planned_memory, &temp_allocator); + + Result method = + program->load_method(*method_name, &memory_manager, nullptr); + if (!method.ok()) { + std::fprintf( + stderr, + "could not load method '%s': 0x%x\n", + *method_name, + static_cast(method.error())); + return 1; + } + + // Inputs are filled with ones so a reference can be computed for any model. + std::vector> host_inputs; + std::vector> sizes; + std::vector> dim_order; + std::vector> strides; +#ifdef RUNNER_HAS_CUDA_DELEGATE + std::vector device_buffers; +#endif + + const size_t num_inputs = method->inputs_size(); + host_inputs.resize(num_inputs); + sizes.resize(num_inputs); + dim_order.resize(num_inputs); + strides.resize(num_inputs); + + for (size_t i = 0; i < num_inputs; ++i) { + Result info = meta->input_tensor_meta(i); + if (!info.ok()) { + continue; // Not a tensor input; leave whatever the program defaults to. + } + // Float32 only, checked rather than assumed. The storage below is sized in + // units of float, so a narrower type would get too few bytes and a wider one + // would be misread. Supporting every type means dtype-aware allocation and + // comparison, which is more than this example needs. + if (info->scalar_type() != executorch::aten::ScalarType::Float) { + std::fprintf(stderr, + "input %zu is not float32, which this runner does not " + "handle\n", + i); + return 1; + } + const auto& shape = info->sizes(); + sizes[i].assign(shape.begin(), shape.end()); + dim_order[i].resize(shape.size()); + strides[i].resize(shape.size()); + int32_t stride = 1; + for (size_t d = shape.size(); d-- > 0;) { + dim_order[i][d] = static_cast(d); + strides[i][d] = stride; + stride *= sizes[i][d]; + } + host_inputs[i].assign(info->nbytes() / sizeof(float), 1.0f); + + void* data = host_inputs[i].data(); +#ifdef RUNNER_HAS_CUDA_DELEGATE + if (gpu_inputs) { + // A memory-planned input is written into the plan's arena with a host + // memcpy, so handing it a device pointer would make the host read device + // memory. This runner allocates the arena on the host, so it takes the + // aliasing path only for a non-planned input. Keeping activations on the + // device instead means giving the runtime a device arena, which is a + // different setup than this example builds. + if (info->is_memory_planned()) { + std::fprintf( + stderr, + "input %zu is memory planned and this runner allocates the plan on " + "the host, so a device pointer would be copied by the host; run " + "without --gpu-inputs for this program\n", + i); + return 1; + } + void* device = nullptr; + if (cudaMalloc(&device, info->nbytes()) != cudaSuccess) { + std::fprintf(stderr, "cudaMalloc failed for input %zu\n", i); + return 1; + } + if (cudaMemcpy( + device, data, info->nbytes(), cudaMemcpyHostToDevice) != + cudaSuccess) { + std::fprintf(stderr, "staging input %zu to the device failed\n", i); + return 1; + } + device_buffers.push_back(device); + data = device; + } +#endif + + auto* impl = method_allocator.allocateInstance< + executorch::runtime::etensor::TensorImpl>(); + new (impl) executorch::runtime::etensor::TensorImpl( + info->scalar_type(), + static_cast(sizes[i].size()), + sizes[i].data(), + data, + dim_order[i].data(), + strides[i].data()); + executorch::runtime::etensor::Tensor tensor(impl); + const Error status = method->set_input(EValue(tensor), i); + if (status != Error::Ok) { + std::fprintf( + stderr, + "set_input(%zu) failed: 0x%x\n", + i, + static_cast(status)); + return 1; + } + } + + const Error status = method->execute(); + if (status != Error::Ok) { + std::fprintf( + stderr, "execute() failed: 0x%x\n", static_cast(status)); + return 1; + } + + std::vector outputs(method->outputs_size()); + if (method->get_outputs(outputs.data(), outputs.size()) != Error::Ok) { + std::fprintf(stderr, "get_outputs() failed\n"); + return 1; + } + + std::vector produced; + for (const EValue& value : outputs) { + if (!value.isTensor()) { + continue; + } + const auto tensor = value.toTensor(); + // Checked for the same reason as the inputs: the read below is in units of + // float, so a narrower type would be read past its end and a wider one + // misinterpreted. + if (tensor.scalar_type() != executorch::aten::ScalarType::Float) { + std::fprintf(stderr, + "an output is not float32, which this runner does not " + "handle\n"); + return 1; + } + const float* data = tensor.const_data_ptr(); + if (data == nullptr) { + std::fprintf(stderr, "an output tensor has no host-readable data\n"); + return 1; + } + produced.insert(produced.end(), data, data + tensor.numel()); + } + + std::printf("produced %zu output values\n", produced.size()); + for (size_t i = 0; i < produced.size() && i < 8; ++i) { + std::printf(" [%zu] %.6f\n", i, produced[i]); + } + +#ifdef RUNNER_HAS_CUDA_DELEGATE + for (void* buffer : device_buffers) { + cudaFree(buffer); + } +#endif + + if (expected_path == nullptr) { + return 0; + } + + const std::vector expected = read_floats(expected_path); + if (expected.empty()) { + std::fprintf(stderr, "could not read expected values from a file\n"); + return 1; + } + if (expected.size() != produced.size()) { + std::fprintf( + stderr, + "expected %zu values but the model produced %zu\n", + expected.size(), + produced.size()); + return 1; + } + double worst = 0.0; + for (size_t i = 0; i < expected.size(); ++i) { + const double difference = + std::abs(static_cast(produced[i]) - expected[i]); + if (difference > worst) { + worst = difference; + } + } + std::printf("largest difference from the reference: %g\n", worst); + if (worst > tolerance) { + std::fprintf( + stderr, "outputs differ by %g, more than the %g allowed\n", worst, + tolerance); + return 1; + } + std::printf("outputs match the reference\n"); + return 0; +} diff --git a/py/torch_tensorrt/executorch/__init__.py b/py/torch_tensorrt/executorch/__init__.py index 123eee846d..5eb5fd9bfd 100644 --- a/py/torch_tensorrt/executorch/__init__.py +++ b/py/torch_tensorrt/executorch/__init__.py @@ -1,5 +1,6 @@ -import importlib -from typing import TYPE_CHECKING, NoReturn +import importlib.util +from pathlib import Path +from typing import NoReturn, Optional, TYPE_CHECKING if TYPE_CHECKING: from executorch.exir import EdgeCompileConfig @@ -41,3 +42,80 @@ def get_edge_compile_config() -> "EdgeCompileConfig": "TensorRTPartitioner", "TensorRTBackend", ] + + +# Handle for the loaded delegate, kept so the load is done once per process. +_runtime_delegate = None + +# Why the load did not happen, kept so a caller that needs the delegate can say what went +# wrong. Without this, a delegate that exists but cannot load surfaces much later as an +# unregistered backend, which describes the symptom and hides the cause. +_runtime_delegate_error: Optional[str] = None + + +def _load_runtime_delegate() -> bool: + """Load the prebuilt TensorRT delegate so it registers itself with ExecuTorch. + + The delegate registers through a static initializer, so it has to be loaded before a + program runs a model that uses it. Returns True when a library was loaded, and False when + this install carries none, which is the case for a build without the delegate. + + Safe to call more than once: the dynamic loader returns the already-loaded library. + """ + global _runtime_delegate, _runtime_delegate_error + if _runtime_delegate is not None: + return True + + package_root = Path(__file__).resolve().parent.parent + # Match any SONAME major so this does not need updating when the major changes. + candidates = sorted(package_root.glob("lib/libexecutorch_backend_tensorrt.so*")) + if not candidates: + _runtime_delegate_error = ( + f"no delegate library found under {package_root / 'lib'}; this package was " + "built without the ExecuTorch delegate" + ) + return False + + import ctypes + + try: + # RTLD_GLOBAL so the delegate can resolve runtime symbols already loaded by the + # ExecuTorch extension, and so anything loaded later can resolve against it. + _runtime_delegate = ctypes.CDLL(str(candidates[0]), mode=ctypes.RTLD_GLOBAL) + except OSError as error: + # A delegate that cannot load is not a reason to make the export path unimportable, + # since exporting needs no delegate. The usual cause is benign: the delegate links the + # ExecuTorch runtime, and a program that only exports has not imported it. + # + # The error is kept rather than discarded. A genuinely broken delegate, a missing CUDA + # or TensorRT dependency or a wrong search path, otherwise surfaces much later as an + # unregistered backend. + _runtime_delegate_error = f"{candidates[0]}: {error}" + return False + return True + + +def require_runtime_delegate() -> None: + """Raise if the prebuilt TensorRT delegate is not loaded. + + Exporting does not need the delegate, so importing this module never fails. A caller about + to run a delegated model does need it, and this reports why it is unavailable instead of + letting the runtime report an unregistered backend later. + + A first attempt can fail for a reason that later stops being true. The delegate links the + ExecuTorch runtime, so importing this module before ExecuTorch leaves that library unfindable, + and the load fails through no fault of the installation. Retrying here means that ordering is + not fatal for the rest of the process. + """ + if _runtime_delegate is not None: + return + if _load_runtime_delegate(): + return + detail = _runtime_delegate_error or "reason unknown" + raise RuntimeError( + "the ExecuTorch TensorRT delegate is not loaded, so a delegated model cannot run: " + f"{detail}" + ) + + +_load_runtime_delegate() diff --git a/setup.py b/setup.py index 2850f73ffa..d7f68123f7 100644 --- a/setup.py +++ b/setup.py @@ -24,7 +24,7 @@ from setuptools.command.develop import develop from setuptools.command.editable_wheel import editable_wheel from setuptools.command.install import install -from torch.utils.cpp_extension import IS_WINDOWS, BuildExtension, CUDAExtension +from torch.utils.cpp_extension import BuildExtension, CUDAExtension, IS_WINDOWS __version__: str = "0.0.0" __cuda_version__: str = "0.0" @@ -98,18 +98,87 @@ def load_dep_info(): RELEASE = False CI_BUILD = False USE_TRT_RTX = False +# Off by default. What a wheel contains should not depend on what happens to be +# installed on the build machine, so the delegate is built only when asked for, and +# then its prerequisites are mandatory. +BUILD_EXECUTORCH_DELEGATE = False -EXECUTORCH_REQUIREMENT = "executorch>=1.3.1" -EXTRAS_REQUIRE = { - "executorch": [EXECUTORCH_REQUIREMENT], - "all": [EXECUTORCH_REQUIREMENT], -} + +# The ExecuTorch release this project's delegate is built and tested against. Written +# here rather than read from the build machine, so the published metadata is a property +# of the source: two builds of the same commit declare the same dependency. Update it +# together with the delegate. +# +# The delegate links the shared runtime through the executorch::runtime target, which +# earlier releases do not provide, so it has to name the release that introduces it. That +# release is still a pre-release, and it is spelled out here because PEP 440 excludes +# pre-releases from == and >= unless the version string says one explicitly. Naming a +# final version that is not published yet would produce metadata no consumer can resolve. +EXECUTORCH_DELEGATE_VERSION = "1.4.0a0" + +# The lowest release that works for a wheel with no delegate. Such a wheel is not bound +# to the newer runtime at all, so it keeps a bound that resolves against what is +# published rather than inheriting the delegate's pre-release pin. +EXECUTORCH_MINIMUM_VERSION = "1.3.1" + + +def _verify_executorch_version() -> None: + """Fail if the installed ExecuTorch is not the pinned one. + + Only called when the delegate is actually being built, since that is what binds the + wheel to one runtime version. There is no promise of a stable C++ ABI across + releases, so a delegate built against a different version than the metadata names + would install cleanly and then fail to load. + """ + try: + from importlib.metadata import version + + installed = version("executorch") + except ModuleNotFoundError: + raise RuntimeError( + "--executorch-delegate needs executorch installed to build against" + ) + # Compare the full public version, including any pre-release marker. Comparing only + # the numeric part would accept 1.4.0a1 or the final 1.4.0 against a 1.4.0a0 pin, and + # those are different binaries from the one the delegate was built against. A local + # build suffix is dropped, since a local build of the pinned release is still that + # release. + installed_release = installed.split("+", 1)[0] + pinned_release = EXECUTORCH_DELEGATE_VERSION.split("+", 1)[0] + if installed_release != pinned_release: + raise RuntimeError( + f"this project pins executorch=={EXECUTORCH_DELEGATE_VERSION} but " + f"{installed} is installed; update EXECUTORCH_DELEGATE_VERSION together " + "with the delegate, or build against the pinned release" + ) + + +def _executorch_requirement(exact: bool) -> str: + """The ExecuTorch requirement for the [executorch] extra. + + Exact only when this wheel carries the delegate, because that is what binds it to one + runtime version. A wheel without the delegate has no such binding, so it keeps a lower + bound against the published releases. + """ + if exact: + return f"executorch=={EXECUTORCH_DELEGATE_VERSION}" + return f"executorch>={EXECUTORCH_MINIMUM_VERSION}" if "--use-rtx" in sys.argv: USE_TRT_RTX = True sys.argv.remove("--use-rtx") +if "--executorch-delegate" in sys.argv: + BUILD_EXECUTORCH_DELEGATE = True + sys.argv.remove("--executorch-delegate") + +EXECUTORCH_REQUIREMENT = _executorch_requirement(BUILD_EXECUTORCH_DELEGATE) +EXTRAS_REQUIRE = { + "executorch": [EXECUTORCH_REQUIREMENT], + "all": [EXECUTORCH_REQUIREMENT], +} + if "--fx-only" in sys.argv: PY_ONLY = True sys.argv.remove("--fx-only") @@ -146,6 +215,18 @@ def load_dep_info(): if use_rtx_env_var == "1" or use_rtx_env_var.lower() == "true": USE_TRT_RTX = True +if BUILD_EXECUTORCH_DELEGATE: + _verify_executorch_version() + +if BUILD_EXECUTORCH_DELEGATE and (PY_ONLY or NO_TS): + # The package_data that carries the delegate is only assembled for a full build, so + # these combinations would compile it and then ship nothing. Fail rather than + # produce a wheel that quietly lacks what was asked for. + raise RuntimeError( + "--executorch-delegate cannot be combined with a Python-only or " + "no-TorchScript build, because those do not package the delegate" + ) + # Distribution name: keep import package as `torch_tensorrt`, but vary project # name so wheels for RTX vs standard TensorRT are distinct. PROJECT_NAME = "torch_tensorrt_rtx" if USE_TRT_RTX else "torch_tensorrt" @@ -178,7 +259,7 @@ def load_dep_info(): RELEASE = True if (gpu_arch_version := os.environ.get("CU_VERSION")) is None: - gpu_arch_version = f"cu{__cuda_version__.replace('.','')}" + gpu_arch_version = f"cu{__cuda_version__.replace('.', '')}" if IS_AARCH64 and (jetpack := os.environ.get("JETPACK_BUILD")) is not None: if jetpack == "1": @@ -323,6 +404,84 @@ def gen_version_file(): f.write('__tensorrt_llm_version__ = "' + __tensorrt_llm_version__ + '"\n') +def build_executorch_delegate(): + """Build and install the ExecuTorch TensorRT delegate into the package. + + This is a CMake project, so it is not part of the Bazel tarball the other + libraries come from. Installing it into the package directory puts the library + under lib/ and its CMake package under lib/cmake, both of which package_data + already collects. + + Skipped when ExecuTorch is not importable, so a build without the optional + dependency keeps working and simply produces a wheel with no delegate. + """ + try: + import executorch + except ImportError: + raise RuntimeError( + "--executorch-delegate needs executorch installed to build against" + ) + + executorch_cmake = ( + Path(executorch.__path__[0]) / "share" / "cmake" / "executorch-config.cmake" + ) + # The file alone is not enough. Released ExecuTorch wheels ship a config that + # only locates the Python extension, with no runtime target to link, so the + # delegate cannot be built against them. Skipping here keeps the wheel + # buildable today and starts producing a delegate once a wheel with the shared + # runtime is available, with no further change needed. + if ( + not executorch_cmake.is_file() + or "executorch::runtime" not in executorch_cmake.read_text() + ): + raise RuntimeError( + "the installed executorch ships no shared runtime target, so the " + "delegate cannot be built against it; a wheel with the C++ runtime is " + "required" + ) + + source = Path(dir_path) / ".." / "cpp" / "src" / "torch_tensorrt" / "executorch" + if not (source / "CMakeLists.txt").is_file(): + raise RuntimeError(f"delegate sources are missing from {source}") + + build = Path(dir_path) / "build" / "executorch_delegate" + prefix = Path(dir_path) / "torch_tensorrt" + configure = [ + "cmake", + "-S", + str(source.resolve()), + "-B", + str(build), + "-DTORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE=ON", + f"-DCMAKE_PREFIX_PATH={executorch_cmake.parent}", + f"-DCMAKE_INSTALL_PREFIX={prefix}", + # Pinned so the library lands where package_data looks, rather than in a + # lib64 directory on distributions that default to it. + "-DCMAKE_INSTALL_LIBDIR=lib", + ] + if os.environ.get("TensorRT_ROOT"): + configure.append(f"-DTensorRT_ROOT={os.environ['TensorRT_ROOT']}") + + print("building the ExecuTorch TensorRT delegate") + if subprocess.run(configure).returncode != 0: + raise RuntimeError("the ExecuTorch TensorRT delegate failed to configure") + subprocess.run(["cmake", "--build", str(build), "-j"], check=True) + subprocess.run(["cmake", "--install", str(build)], check=True) + + # Assert what the wheel will actually carry. Without this a configuration that + # produces nothing still reports success, and the wheel ships without the + # delegate it advertises. + library = sorted((prefix / "lib").glob("libexecutorch_backend_tensorrt.so*")) + if not library: + raise RuntimeError(f"the delegate library was not installed into {prefix}") + exported = prefix / "lib" / "cmake" / "torch_tensorrt_executorch" + if not list(exported.glob("*.cmake")): + raise RuntimeError( + f"the delegate CMake package was not installed into {exported}" + ) + print(f"packaged the ExecuTorch TensorRT delegate: {library[0].name}") + + def copy_libtorchtrt(multilinux=False, rt_only=False): if not os.path.exists(dir_path + "/torch_tensorrt/lib"): os.makedirs(dir_path + "/torch_tensorrt/lib") @@ -377,6 +536,9 @@ def finalize_options(self): self.root_is_pure = False def run(self): + # Independent of the Bazel build: this needs only an installed ExecuTorch. + if BUILD_EXECUTORCH_DELEGATE: + build_executorch_delegate() if not PY_ONLY: build_libtorchtrt_cxx11_abi(develop=True, rt_only=NO_TS) copy_libtorchtrt(rt_only=NO_TS) @@ -397,6 +559,9 @@ def finalize_options(self): self.root_is_pure = False def run(self): + # Independent of the Bazel build: this needs only an installed ExecuTorch. + if BUILD_EXECUTORCH_DELEGATE: + build_executorch_delegate() if not PY_ONLY: build_libtorchtrt_cxx11_abi(develop=False, rt_only=NO_TS) copy_libtorchtrt(rt_only=NO_TS) @@ -423,6 +588,9 @@ def run(self): self.distribution.metadata.name = PROJECT_NAME except Exception: pass + # Independent of the Bazel build: this needs only an installed ExecuTorch. + if BUILD_EXECUTORCH_DELEGATE: + build_executorch_delegate() if not PY_ONLY: build_libtorchtrt_cxx11_abi(develop=False, rt_only=NO_TS) copy_libtorchtrt(rt_only=NO_TS) @@ -613,6 +781,12 @@ def run(self): package_data = {} executorch_header_package_data = ["include/torch_tensorrt/executorch/*.h"] +# The delegate is built by CMake rather than Bazel, so it is copied into the +# package separately. Its CMake package has to ship too, otherwise a C++ +# application can find the library but not the target that knows how to link it. +executorch_delegate_package_data = [ + "lib/cmake/torch_tensorrt_executorch/*.cmake", +] if not (PY_ONLY or NO_TS): tensorrt_x86_64_external_dir = ( @@ -792,6 +966,7 @@ def run(self): "torch_tensorrt": [ "include/torch_tensorrt/*.h", *executorch_header_package_data, + *executorch_delegate_package_data, "include/torch_tensorrt/core/*.h", "include/torch_tensorrt/core/conversion/*.h", "include/torch_tensorrt/core/conversion/conversionctx/*.h", @@ -822,6 +997,7 @@ def run(self): "torch_tensorrt": [ "include/torch_tensorrt/*.h", *executorch_header_package_data, + *executorch_delegate_package_data, "include/torch_tensorrt/core/*.h", "include/torch_tensorrt/core/runtime/*.h", "lib/*", diff --git a/tests/py/dynamo/executorch/test_wheel_e2e.py b/tests/py/dynamo/executorch/test_wheel_e2e.py new file mode 100644 index 0000000000..c96531dfc8 --- /dev/null +++ b/tests/py/dynamo/executorch/test_wheel_e2e.py @@ -0,0 +1,361 @@ +"""End-to-end test: export with TensorRT, then run from C++ against wheels only. + +The other tests in this directory stop at the serialized program, which is what +they say they do: they assert the right delegates are present but never load or +run anything. This one goes further and is the only test that proves the shipped +C++ pieces are usable: + +1. Export a model where TensorRT actually claims work, and fail if it claimed + nothing. Without that check the rest of the test proves nothing, because a + program with zero TensorRT delegates still loads and runs fine. +2. Build a C++ application against INSTALLED packages, with no source checkout, + using find_package for both the runtime and the delegate. +3. Run it and compare against a reference computed in Python. +4. Repeat with the input tensors allocated in device memory, because that is how + an accelerator application feeds data. The CUDA delegate copies device to + device and never stages through the host, so it expects tensors that are + already resident. +""" + +import importlib.util +import os +import shutil +import subprocess +import sys +from pathlib import Path + +import pytest + +pytest.importorskip("executorch.exir") + +import torch # noqa: E402 + +# Importing this pulls in the compiled TensorRT runtime, which is absent in a +# collection-only or CPU-only environment. Skipping rather than failing keeps the +# rest of the suite collectable there. +torch_tensorrt = pytest.importorskip("torch_tensorrt") + +_RUNNER_SOURCE_DIR = ( + Path(__file__).resolve().parents[4] / "examples" / "executorch_wheel_runner" +) + + +def _installed_executorch_cmake_dir(): + """Where the installed ExecuTorch package keeps its CMake config. + + None unless the config actually offers the shared runtime target. Released wheels + ship the file without it, and the example needs that target, so returning the + directory anyway would let the test configure and fail rather than skip. + """ + spec = importlib.util.find_spec("executorch") + if spec is None or not spec.submodule_search_locations: + return None + root = Path(list(spec.submodule_search_locations)[0]) + config = root / "share" / "cmake" / "executorch-config.cmake" + if not config.is_file(): + return None + if "executorch::runtime" not in config.read_text(): + return None + return config.parent + + +_CMAKE_DIR = _installed_executorch_cmake_dir() + +_REPO_ROOT = Path(__file__).resolve().parents[4] +_DELEGATE_SOURCE_DIR = _REPO_ROOT / "cpp" / "src" / "torch_tensorrt" / "executorch" + + +@pytest.fixture(scope="module") +def delegate_prefix(tmp_path_factory): + """Build and install the TensorRT delegate, and return its CMake prefix. + + Without this the runner is configured against ExecuTorch alone, so + find_package(torch_tensorrt_executorch) finds nothing and the application is + built without the delegate. A program that needs TensorRTBackend would then run + against a runner that never linked it, which cannot work and would not say why. + """ + if _CMAKE_DIR is None or shutil.which("cmake") is None: + pytest.skip("needs an ExecuTorch CMake package and cmake") + if not (_DELEGATE_SOURCE_DIR / "CMakeLists.txt").is_file(): + pytest.skip("needs the delegate sources, which a wheel-only install lacks") + + root = tmp_path_factory.mktemp("delegate") + prefix = root / "install" + configure = [ + "cmake", + "-S", + str(_DELEGATE_SOURCE_DIR), + "-B", + str(root / "build"), + "-DTORCHTRT_EXECUTORCH_BUILD_SHARED_DELEGATE=ON", + f"-DCMAKE_PREFIX_PATH={_CMAKE_DIR}", + f"-DCMAKE_INSTALL_PREFIX={prefix}", + # Pinned so the check below does not have to guess between lib and lib64. + "-DCMAKE_INSTALL_LIBDIR=lib", + ] + if os.environ.get("TensorRT_ROOT"): + configure.append(f"-DTensorRT_ROOT={os.environ['TensorRT_ROOT']}") + + result = subprocess.run(configure, capture_output=True, text=True, check=False) + if result.returncode != 0: + pytest.skip(f"the delegate does not configure here: {result.stderr[-400:]}") + subprocess.run( + ["cmake", "--build", str(root / "build"), "-j"], + capture_output=True, + text=True, + check=True, + ) + subprocess.run(["cmake", "--install", str(root / "build")], check=True) + + installed = list(prefix.rglob("libexecutorch_backend_tensorrt.so*")) + assert installed, f"the delegate installed no shared library under {prefix}" + return prefix + + +requires_wheel = pytest.mark.skipif( + _CMAKE_DIR is None, + reason="needs an ExecuTorch install that ships its CMake package", +) +requires_cmake = pytest.mark.skipif( + shutil.which("cmake") is None, reason="needs cmake to build the C++ application" +) +requires_cuda = pytest.mark.skipif( + not torch.cuda.is_available(), reason="needs a CUDA device" +) + + +class _Model(torch.nn.Module): + """Small graph that TensorRT can take in full.""" + + def forward(self, x): + return torch.tanh(x * 2.0 + 1.0) + + +def _delegate_ids(pte_path: Path): + """Backend ids of every delegate in the serialized program, in order.""" + from executorch.exir._serialize._program import deserialize_pte_binary + + program = deserialize_pte_binary(pte_path.read_bytes()).program + return [ + delegate.id for plan in program.execution_plan for delegate in plan.delegates + ] + + +def _export(model, example_input, destination: Path) -> Path: + exported = torch.export.export(model, (example_input,)) + compiled = torch_tensorrt.dynamo.compile( + exported, + arg_inputs=[ + torch_tensorrt.Input( + shape=tuple(example_input.shape), dtype=example_input.dtype + ) + ], + min_block_size=1, + ) + torch_tensorrt.save( + compiled, + str(destination), + output_format="executorch", + arg_inputs=(example_input,), + retrace=False, + ) + return destination + + +def _skip_on_engine_version_mismatch(result, caplog_text=""): + """Skip when the engine and the TensorRT runtime are different versions. + + A serialized engine only loads on the TensorRT version that produced it. When + the export environment and the linked runtime disagree, every delegated run + fails for that reason alone, which says nothing about the packaging this test + covers. Any other failure is still a real failure. + + The mismatch can surface either from the application or from TensorRT's own + logger during export, so both sources are checked. A delegated program whose + operators are all missing is the same situation seen from the runtime side: + the engine never deserialized, so the delegate produced no kernels. + """ + combined = result.stdout + result.stderr + caplog_text + markers = ( + "Serialized Engine Version", + "version mismatch", + "Version tag does not match", + ) + if any(marker in combined for marker in markers): + pytest.skip( + "the serialized engine and the linked TensorRT runtime are different " + "versions, so a delegated run cannot succeed in this environment" + ) + + +def _build_runner(build_dir: Path, delegate_prefix: Path = None) -> Path: + prefixes = [str(_CMAKE_DIR)] + if delegate_prefix is not None: + prefixes.append(str(delegate_prefix)) + subprocess.run( + [ + "cmake", + "-S", + str(_RUNNER_SOURCE_DIR), + "-B", + str(build_dir), + "-DCMAKE_PREFIX_PATH={}".format(";".join(prefixes)), + ], + check=True, + ) + subprocess.run(["cmake", "--build", str(build_dir), "-j"], check=True) + runner = build_dir / "executorch_wheel_runner" + assert runner.is_file(), "the C++ application did not get built" + if delegate_prefix is not None: + # The delegate registers itself from a static initializer, so a runner that + # merely built is not enough: the library has to still be on the link line. + needed = subprocess.run( + ["readelf", "-d", str(runner)], capture_output=True, text=True, check=False + ).stdout + assert "libexecutorch_backend_tensorrt" in needed, ( + "the runner built without the TensorRT delegate, so a delegated program " + "cannot run against it" + ) + return runner + + +def _write_reference(model, example_input, path: Path) -> None: + with torch.no_grad(): + reference = model(example_input) + values = reference.detach().cpu().flatten().tolist() + path.write_text(" ".join(f"{value:.6f}" for value in values)) + + +@requires_cuda +@requires_wheel +@requires_cmake +def test_tensorrt_partitions_run_from_cpp(tmp_path, delegate_prefix, caplog): + """A TensorRT-delegated program runs from C++ and matches a reference.""" + model = _Model().eval().cuda() + example_input = torch.ones((2, 3, 4, 4)).cuda() + + pte = _export(model, example_input, tmp_path / "model.pte") + assert pte.is_file() + + # Without this the test would pass even if TensorRT claimed nothing at all. + delegates = _delegate_ids(pte) + assert delegates.count("TensorRTBackend") > 0, ( + f"TensorRT claimed no part of the graph, so this proves nothing about the " + f"delegate; delegates were {delegates}" + ) + + reference = tmp_path / "expected.txt" + # The runner fills inputs with ones, so the reference uses the same input. + _write_reference(model, torch.ones_like(example_input), reference) + + runner = _build_runner(tmp_path / "build", delegate_prefix) + result = subprocess.run( + [ + str(runner), + "--model", + str(pte), + "--expected", + str(reference), + "--tolerance", + "1e-3", + ], + capture_output=True, + text=True, + ) + sys.stderr.write(result.stderr) + _skip_on_engine_version_mismatch(result, caplog.text) + assert result.returncode == 0, "the C++ application did not match the reference" + assert "outputs match the reference" in result.stdout + + +@requires_cuda +@requires_wheel +@requires_cmake +def test_device_resident_inputs(tmp_path, delegate_prefix, caplog): + """The same program runs when its inputs start in device memory. + + An accelerator application allocates its tensors on the device and hands those + pointers to the runtime. This checks that path rather than the host one. + """ + model = _Model().eval().cuda() + example_input = torch.ones((2, 3, 4, 4)).cuda() + + pte = _export(model, example_input, tmp_path / "model.pte") + delegates = _delegate_ids(pte) + assert delegates.count("TensorRTBackend") > 0, ( + f"TensorRT claimed no part of the graph; delegates were {delegates}" + ) + + reference = tmp_path / "expected.txt" + _write_reference(model, torch.ones_like(example_input), reference) + + runner = _build_runner(tmp_path / "build", delegate_prefix) + result = subprocess.run( + [ + str(runner), + "--model", + str(pte), + "--expected", + str(reference), + "--tolerance", + "1e-3", + "--gpu-inputs", + ], + capture_output=True, + text=True, + ) + sys.stderr.write(result.stderr) + if result.returncode == 2: + pytest.skip("the installed wheel has no CUDA delegate, so no device inputs") + _skip_on_engine_version_mismatch(result, caplog.text) + # The runner refuses device pointers for a memory-planned input, because the + # runtime would copy such an input into the plan with a host memcpy. That is + # the runner behaving correctly for this program, not a packaging failure. + if "is memory planned" in result.stdout + result.stderr: + pytest.skip( + "this program's inputs are memory planned, so the runtime copies them " + "on the host and device-resident inputs do not apply" + ) + assert result.returncode == 0, "device-resident inputs did not match the reference" + assert "outputs match the reference" in result.stdout + + +@requires_wheel +@requires_cmake +def test_runner_uses_no_source_checkout(tmp_path): + """The application builds from installed packages alone. + + This is the property the whole prebuilt-library effort exists to provide, so it + is worth asserting rather than assuming: configuring must not need an + ExecuTorch source tree anywhere. + """ + build_dir = tmp_path / "build" + configure = subprocess.run( + [ + "cmake", + "-S", + str(_RUNNER_SOURCE_DIR), + "-B", + str(build_dir), + f"-DCMAKE_PREFIX_PATH={_CMAKE_DIR}", + ], + capture_output=True, + text=True, + ) + assert configure.returncode == 0, configure.stderr + # The example's own output, not ExecuTorch's. Matching another project's log wording + # would break whenever that project rephrases a message, while this line is emitted + # by the code under test and only when the runtime target actually resolved. + assert "linking executorch::" in configure.stdout, ( + "the example did not report linking any ExecuTorch component, so the package " + f"config did not provide the targets it needs: {configure.stdout[-400:]}" + ) + + # A source build would have configured ExecuTorch itself, leaving its cache + # behind. Only the application's own cache should exist. + caches = list(build_dir.rglob("CMakeCache.txt")) + assert len(caches) == 1, ( + f"expected only the application's own CMake cache, found {caches}. More than " + f"one means an ExecuTorch source tree was configured as a subproject." + ) + assert not (build_dir / "executorch").exists()