Load AOTI plugins into hosts that don't share their C++ runtime

Summary:
An AOTInductor plugin is compiled separately from the program that loads it, so the two don't have to share a C++ standard library. Two places in the CUDA backend assume they do.

- **Constants pass through a `std::unordered_map`.** Cross-method weight sharing, mutable-state sessions and the CUDA graph host-constant check read a method's constants with `AOTInductorModelContainerExtractConstantsMap`, which hands the plugin a pointer to the host's map to fill in. That's only safe when both sides use the same standard library. It isn't safe for a Windows plugin cross-compiled with MinGW and loaded by an MSVC host, or a plugin loaded by a libc++ host. PyTorch's AOTI interface marks that call deprecated in favour of `AOTInductorModelContainerExtractConstantsMapEntries`, which returns a container-owned array of `{name, handle}` C structs. The backend now reads constants only through that call, so no C++ container crosses the plugin boundary.
- **The plugin's C++ runtime has to already be in the process.** On Linux an x86 plugin references libstdc++ (`std::cerr`, exception handling, `operator new`) without listing `libstdc++.so.6` as a dependency. Loaded from a process that doesn't link libstdc++, `dlopen` fails with `undefined symbol: _ZSt4cerr`. When the host does link libstdc++, the plugin loads only if the host's dynamic symbol table happens to include every libstdc++ symbol the plugin uses. `load_library` now opens `libstdc++.so.6` with `RTLD_GLOBAL` once before the first plugin, and returns `AccessFailed` if it can't, since such a plugin can't load anyway. The preload adds libstdc++ to the global lookup scope after the host, so symbols the host already defines still win.

Plugins that don't export the entries call still run, with reduced features:
- each method loads its own weights from the blob, without cross-method sharing;
- mutable-state sessions report `NotSupported`;
- CUDA graphs stay off, since their constants can't be checked for host memory.

Malformed results are never dereferenced: a null array with a nonzero count, a null name or handle, or an extraction error is rejected with `InvalidProgram`, or keeps graphs off.

`test_cuda_mutable_state` gains cases for the graph check (no entries call, no constants, a host constant, a device constant, malformed results) and for mutable-state extraction failures. It also builds outside fbcode as `test_cuda_mutable_state_native`.

Differential Revision: D129654474

Author: mattiasbrand

backends/aoti/aoti_delegate_handle.h

@@ -114,6 +114,14 @@ using AOTInductorModelContainerExtractConstantsMapFunc = AOTIRuntimeError (*)(
     AOTInductorConstantMapHandle constant_map_handle,
     bool use_inactive);
 
+// Extracts borrowed entries through a flat C ABI; the container owns the array.
+using AOTInductorModelContainerExtractConstantsMapEntriesFunc =
+    AOTIRuntimeError (*)(
+        AOTInductorModelContainerHandle container_handle,
+        const AOTInductorConstantMapEntry** entries,
+        size_t* num_entries,
+        bool use_inactive);
+
 // Updates the container's constants with user-managed tensor handles.
 // DLL-boundary safe — uses a flat C array instead of std::unordered_map.
 using AOTInductorModelContainerUpdateUserManagedConstantBufferPairsFunc =
@@ -146,6 +154,8 @@ struct AOTIDelegateHandle {
   AOTInductorModelContainerGetConstantNameFunc get_constant_name;
   AOTInductorModelContainerGetConstantOriginalFQNFunc get_constant_original_fqn;
   AOTInductorModelContainerExtractConstantsMapFunc extract_constants_map;
+  AOTInductorModelContainerExtractConstantsMapEntriesFunc
+      extract_constants_map_entries{nullptr};
   AOTInductorModelContainerUpdateUserManagedConstantBufferPairsFunc
       update_user_managed_constant_buffer_pairs;
 };

backends/cuda/runtime/cuda_backend.cpp

@@ -233,6 +233,9 @@ class ET_EXPERIMENTAL CudaBackend final
   Error load_function_pointers_into_handle(
       void* so_handle,
       cuda::CudaDelegateHandle* handle) const {
+    if (handle == nullptr) {
+      return Error::InvalidArgument;
+    }
 #define LOAD_SYMBOL(member, name)                                    \
   do {                                                               \
     auto symbol_res = get_function(so_handle, #name);                \
@@ -299,6 +302,9 @@ class ET_EXPERIMENTAL CudaBackend final
         AOTInductorModelContainerGetConstantFromFolded);
     LOAD_OPTIONAL_SYMBOL(
         extract_constants_map, AOTInductorModelContainerExtractConstantsMap);
+    LOAD_OPTIONAL_SYMBOL(
+        extract_constants_map_entries,
+        AOTInductorModelContainerExtractConstantsMapEntries);
     LOAD_OPTIONAL_SYMBOL(
         update_user_managed_constant_buffer_pairs,
         AOTInductorModelContainerUpdateUserManagedConstantBufferPairs);
@@ -628,7 +634,8 @@ class ET_EXPERIMENTAL CudaBackend final
     // Initialize CUDA graph state if enabled for this method. Not on a device
     // without memory pools, where the method runs without a graph.
     if (should_use_cuda_graph_for_method(method_name)) {
-      const std::string host_constants = describe_host_constants(handle);
+      const std::string host_constants =
+          cuda::describe_host_constants_for_graph(*handle);
       if (!host_constants.empty()) {
         // A graph records only the device work of the capture call. Host code,
         // such as the choice of a torch.cond branch on a flag in host memory,
@@ -948,29 +955,6 @@ class ET_EXPERIMENTAL CudaBackend final
   }
 
  private:
-  // Names a constant of the program in host memory, or says that the
-  // constants could not be listed, since unknown counts as host memory.
-  // Empty when every constant is on the device.
-  static std::string describe_host_constants(cuda::CudaDelegateHandle* handle) {
-    const char* unlisted = "its constants could not be listed";
-    if (handle->extract_constants_map == nullptr) {
-      return unlisted;
-    }
-    std::unordered_map<std::string, AtenTensorHandle> constants;
-    if (handle->extract_constants_map(
-            handle->container_handle,
-            reinterpret_cast<AOTInductorConstantMapHandle>(&constants),
-            /*use_inactive=*/false) != Error::Ok) {
-      return unlisted;
-    }
-    for (const auto& [fqn, tensor] : constants) {
-      if (reinterpret_cast<SlimTensor*>(tensor)->is_cpu()) {
-        return "constant '" + fqn + "' is in host memory";
-      }
-    }
-    return "";
-  }
-
   // Runs the program once. On the capture step it captures the CUDA graph
   // instead, then launches it once, and sets graph_launched once that launch
   // is queued.
@@ -1513,9 +1497,13 @@ class ET_EXPERIMENTAL CudaBackend final
       const NamedDataMap* named_data_map,
       const std::string& method_name,
       const std::string& weights_blob_key) const {
+    if (handle == nullptr || named_data_map == nullptr) {
+      return Error::InvalidArgument;
+    }
     // Check if the required APIs are available
     if (!handle->get_num_constants || !handle->get_constant_name ||
-        !handle->get_constant_original_fqn || !handle->extract_constants_map ||
+        !handle->get_constant_original_fqn ||
+        !handle->extract_constants_map_entries ||
         !handle->update_user_managed_constant_buffer_pairs) {
       // Fall back to the legacy path
       return load_constants_legacy(handle, named_data_map, weights_blob_key);
@@ -1581,6 +1569,13 @@ class ET_EXPERIMENTAL CudaBackend final
     std::unordered_map<std::string, AtenTensorHandle> extracted_map;
 
     if (!uncached_fqns.empty()) {
+      const auto extract = handle->extract_constants_map_entries;
+      if (extract == nullptr) {
+        ET_LOG(
+            Error,
+            "CUDA blob weight sharing requires C-ABI constants extraction");
+        return Error::NotSupported;
+      }
       // Need to load from blob — use update_constants_from_blob for all,
       // then extract the new constants into the cache.
       auto buffer_res = named_data_map->get_data(weights_blob_key.c_str());
@@ -1606,14 +1601,33 @@ class ET_EXPERIMENTAL CudaBackend final
       ET_CUDA_CHECK_OR_RETURN_ERROR(cudaDeviceSynchronize());
       buffer_res->Free();
 
-      // Extract all constants from the freshly-loaded container.
+      const AOTInductorConstantMapEntry* entries = nullptr;
+      size_t num_entries = 0;
       ET_CHECK_OK_OR_RETURN_ERROR(
-          handle->extract_constants_map(
+          extract(
               handle->container_handle,
-              reinterpret_cast<AOTInductorConstantMapHandle>(&extracted_map),
+              &entries,
+              &num_entries,
               /*use_inactive=*/false),
           "Failed to extract constants from '%s'",
           method_name.c_str());
+      if (entries == nullptr && num_entries != 0) {
+        ET_LOG(
+            Error, "CUDA constants extraction returned a null entries array");
+        return Error::InvalidProgram;
+      }
+      if (entries != nullptr) {
+        for (size_t i = 0; i < num_entries; ++i) {
+          const auto entry = entries[i];
+          if (entry.name == nullptr || entry.handle == nullptr) {
+            ET_LOG(
+                Error,
+                "CUDA constants extraction returned a null name or tensor handle");
+            return Error::InvalidProgram;
+          }
+          extracted_map.emplace(entry.name, entry.handle);
+        }
+      }
 
       // Validate cache hits against the freshly-extracted tensors, and
       // populate the cache with newly-loaded entries.

backends/cuda/runtime/cuda_delegate_handle.h

@@ -356,6 +356,35 @@ struct CudaDelegateHandle : public aoti::AOTIDelegateHandle {
   std::unordered_map<std::string, std::vector<int64_t>> offgraph_kv_sizes;
 };
 
+// An unknown or host-resident constant prevents safe CUDA graph capture.
+// Compiled plugins return flat entries so C++ containers stay host-owned.
+inline std::string describe_host_constants_for_graph(
+    const CudaDelegateHandle& handle) {
+  const char* unlisted = "its constants could not be listed";
+  if (handle.extract_constants_map_entries == nullptr) {
+    return unlisted;
+  }
+  const aoti::AOTInductorConstantMapEntry* entries = nullptr;
+  size_t count = 0;
+  if (handle.extract_constants_map_entries(
+          handle.container_handle, &entries, &count, /*use_inactive=*/false) !=
+          executorch::runtime::Error::Ok ||
+      (entries == nullptr && count != 0)) {
+    return unlisted;
+  }
+  for (size_t i = 0; i < count; ++i) {
+    if (entries[i].name == nullptr || entries[i].handle == nullptr) {
+      return unlisted;
+    }
+    if (reinterpret_cast<aoti::slim::SlimTensor*>(entries[i].handle)
+            ->is_cpu()) {
+      return std::string("constant '") + entries[i].name +
+          "' is in host memory";
+    }
+  }
+  return "";
+}
+
 } // namespace cuda
 } // namespace backends
 } // namespace executorch

backends/cuda/runtime/cuda_mutable_state.cpp

@@ -101,8 +101,8 @@ thread_local MutableStateContext tl_active_ctx = kInvalidMutableContext;
 thread_local int tl_active_token = kNoMutableSession;
 
 bool handle_has_symbols(CudaDelegateHandle* h) {
-  return h->get_num_constants && h->get_constant_name &&
-      h->get_constant_original_fqn && h->extract_constants_map &&
+  return h != nullptr && h->get_num_constants && h->get_constant_name &&
+      h->get_constant_original_fqn && h->extract_constants_map_entries &&
       h->update_user_managed_constant_buffer_pairs;
 }
 
@@ -255,6 +255,13 @@ Error validate_coverage_locked(Context& c) {
 // Captures descriptors and initial templates while the container still owns its
 // default mutable buffers.
 Error build_descriptors(Context& c, CudaDelegateHandle* h) {
+  if (h == nullptr) {
+    return Error::InvalidArgument;
+  }
+  const auto extract = h->extract_constants_map_entries;
+  if (extract == nullptr) {
+    return Error::NotSupported;
+  }
   auto container = h->container_handle;
 
   size_t n = 0;
@@ -278,12 +285,27 @@ Error build_descriptors(Context& c, CudaDelegateHandle* h) {
   }
 
   std::unordered_map<std::string, aoti::AtenTensorHandle> extracted;
+  const aoti::AOTInductorConstantMapEntry* entries = nullptr;
+  size_t num_entries = 0;
   ET_CHECK_OK_OR_RETURN_ERROR(
-      h->extract_constants_map(
-          container,
-          reinterpret_cast<aoti::AOTInductorConstantMapHandle>(&extracted),
-          /*use_inactive=*/false),
-      "mutable_state: extract_constants_map failed");
+      extract(container, &entries, &num_entries, /*use_inactive=*/false),
+      "mutable_state: extract_constants_map_entries failed");
+  if (entries == nullptr && num_entries != 0) {
+    ET_LOG(
+        Error,
+        "mutable_state: constants extraction returned a null entries array");
+    return Error::InvalidProgram;
+  }
+  if (entries != nullptr) {
+    for (size_t i = 0; i < num_entries; ++i) {
+      const auto entry = entries[i];
+      if (entry.name == nullptr || entry.handle == nullptr) {
+        ET_LOG(Error, "mutable_state: null constant name or tensor handle");
+        return Error::InvalidProgram;
+      }
+      extracted.emplace(entry.name, entry.handle);
+    }
+  }
 
   auto& table = c.desc[h];
   for (const auto& fqn : c.fqns) {

backends/cuda/runtime/platform/platform.cpp

@@ -76,6 +76,20 @@ executorch::runtime::Result<void*> load_library(
   }
 
 #else
+#if defined(__linux__)
+  // AOTI plugins use a C ABI but can leave GNU C++ runtime references
+  // unresolved. Native hosts may use a different C++ standard library.
+  static void* const gnu_cpp_runtime = []() {
+    void* handle = dlopen("libstdc++.so.6", RTLD_NOW | RTLD_GLOBAL);
+    if (handle == nullptr) {
+      ET_LOG(Error, "Failed to load the AOTI GNU C++ runtime: %s", dlerror());
+    }
+    return handle;
+  }();
+  if (gnu_cpp_runtime == nullptr) {
+    return executorch::runtime::Error::AccessFailed;
+  }
+#endif
   // Before loading the delegate .so, we need to ensure symbols from the current
   // process (e.g. this backend library) are globally visible. Python loads
   // modules with RTLD_LOCAL by default, so we re-open the current module with

backends/cuda/runtime/targets.bzl

@@ -164,6 +164,31 @@ def define_common_targets(is_fbcode = False):
     )
 
     if not is_fbcode:
+        runtime.genrule(
+            name = "test_cuda_runtime_libraries",
+            out = "libs",
+            platforms = [CXX],
+            compatible_with = ["ovr_config//cuda:has_cuda"],
+            cmd = """
+                mkdir -p "$OUT"
+                cp $(location fbsource//third-party/cuda:cublas) "$OUT/"
+                cp $(location fbsource//third-party/cuda:cublasLt) "$OUT/"
+                cp $(location fbsource//third-party/cuda:cudart) "$OUT/"
+                cp $(location fbsource//third-party/cuda:curand) "$OUT/"
+            """,
+        )
+        runtime.cxx_test(
+            name = "test_cuda_mutable_state_native",
+            srcs = ["test/test_cuda_mutable_state.cpp"],
+            env = {"LD_LIBRARY_PATH": "$(location :test_cuda_runtime_libraries)"},
+            platforms = [CXX],
+            compatible_with = ["ovr_config//cuda:has_cuda"],
+            deps = [
+                ":cuda_backend",
+                "//executorch/backends/aoti/slim/factory:from_blob",
+                "fbsource//third-party/googletest:gtest_main",
+            ],
+        )
         return
 
     cpp_unittest(