Improve CUDA backend error handling and add dual-method runner fallback

- cuda_backend.cpp: Replace ET_CHECK_OK_OR_RETURN_ERROR with explicit error
  handling + cudaDeviceSynchronize after weight transfer, add logging for
  missing weights_blob
- main.cpp: Support single "forward" method fallback when prefill/decode
  not available, use prefill_method variable, remove debug printf

Author: gasoonjia

backends/cuda/runtime/cuda_backend.cpp

@@ -372,9 +372,18 @@ class ET_EXPERIMENTAL CudaBackend final
       const void* weights_blob = buffer_res->data();
       // Feed the weights blob into the container. Under the hood it's copying
       // weights, so we should free the buffer immediately.
-      ET_CHECK_OK_OR_RETURN_ERROR(handle->update_constants_from_blob(
-          handle->container_handle, static_cast<const uint8_t*>(weights_blob)));
+      auto update_err = handle->update_constants_from_blob(
+          handle->container_handle, static_cast<const uint8_t*>(weights_blob));
+      if (update_err != Error::Ok) {
+        ET_LOG(Error, "update_constants_from_blob failed");
+        return update_err;
+      }
+      // Ensure all weight transfers are complete before execution
+      cudaDeviceSynchronize();
       buffer_res->Free();
+    } else {
+      ET_LOG(Info, "weights_blob '%s' not found or update fn is null",
+          weights_blob_key.c_str());
     }
 
     // Use shared CUDA stream if enabled via options, otherwise create one.

examples/models/qwen3_5_moe/main.cpp

@@ -86,16 +86,27 @@ int main(int argc, char** argv) {
 
   printf("Loading methods...\n");
 
-  // Load both methods
+  // Try loading both methods; fall back to single "forward" method
+  bool dual_method = true;
+  std::string prefill_method = "prefill";
   auto err = module->load_method("prefill");
   if (err != Error::Ok) {
-    ET_LOG(Error, "Failed to load prefill method");
-    return 1;
+    // Try "forward" for single-method export
+    err = module->load_method("forward");
+    if (err != Error::Ok) {
+      ET_LOG(Error, "Failed to load prefill/forward method");
+      return 1;
+    }
+    prefill_method = "forward";
+    dual_method = false;
+    printf("Using single-method mode (forward)\n");
   }
-  err = module->load_method("decode");
-  if (err != Error::Ok) {
-    ET_LOG(Error, "Failed to load decode method");
-    return 1;
+  if (dual_method) {
+    err = module->load_method("decode");
+    if (err != Error::Ok) {
+      ET_LOG(Error, "Failed to load decode method");
+      return 1;
+    }
   }
 
   // Get EOS ids
@@ -138,7 +149,7 @@ int main(int argc, char** argv) {
   prefill_inputs.push_back(tokens_tensor);
   prefill_inputs.push_back(pos_tensor);
 
-  auto prefill_result = module->execute("prefill", prefill_inputs);
+  auto prefill_result = module->execute(prefill_method, prefill_inputs);
   if (prefill_result.error() != Error::Ok) {
     ET_LOG(Error, "Prefill failed");
     return 1;
@@ -165,6 +176,11 @@ int main(int argc, char** argv) {
   // decode method, which may run on a different CUDA stream.
   cudaDeviceSynchronize();
 
+  if (!dual_method) {
+    printf("Single-method mode: skipping decode\n");
+    return 0;
+  }
+
   // ---------------------------------------------------------------
   // Decode — generate tokens one at a time
   // ---------------------------------------------------------------