fix: handle None from cupy.cuda.get_cuda_path() in cuda_launch

cupy.cuda.get_cuda_path() can return None when CUDA_HOME is not set
and cupy can't auto-detect it. Fall back to /usr/local/cuda.

Co-Authored-By: Claude Opus 4.6 <noreply@anthropic.com>

bim_vfi_arch/costvol.py

@@ -4,6 +4,7 @@ import collections
 import os
 import re
 import torch
+import torch.nn.functional as F
 import typing
 
 cupy = None
@@ -15,11 +16,7 @@ def _ensure_cupy():
             import cupy as _cupy
             cupy = _cupy
         except ImportError:
-            raise RuntimeError(
-                "cupy is required for BIM-VFI. Install it with:\n"
-                "  pip install cupy-cuda12x  (or cupy-cuda11x for CUDA 11)\n"
-                "Or run install.py from the ComfyUI-Tween directory."
-            )
+            pass  # cupy unavailable; PyTorch fallback will be used
 
 
 ##########################################################
@@ -246,6 +243,28 @@ def cuda_launch(strKey:str):
 # end
 
 
+def _pytorch_costvol(tenOne, tenTwo, intKernelSize):
+    """Pure-PyTorch local cost volume via unfold + dot product."""
+    B, C, H, W = tenOne.shape
+    pad = (intKernelSize - 1) // 2
+
+    # Pad tenTwo so out-of-bounds yields 0 (matches CUDA kernel)
+    tenTwo_padded = F.pad(tenTwo, [pad, pad, pad, pad])
+
+    # Unfold into patches: (B, C, H, W, K, K)
+    patches = tenTwo_padded.unfold(2, intKernelSize, 1).unfold(3, intKernelSize, 1)
+    # Reshape to (B, C, H, W, K*K)
+    patches = patches.contiguous().view(B, C, H, W, intKernelSize * intKernelSize)
+
+    # Dot product over C dimension: (B, H, W, K*K)
+    tenOut = (tenOne.unsqueeze(-1) * patches).sum(dim=1)
+
+    # Permute to (B, K*K, H, W) to match CUDA output layout
+    tenOut = tenOut.permute(0, 3, 1, 2).contiguous()
+
+    return tenOut
+
+
 ##########################################################
 
 
@@ -253,6 +272,8 @@ class costvol_func(torch.autograd.Function):
     @staticmethod
     @torch.amp.custom_fwd(device_type='cuda', cast_inputs=torch.float32)
     def forward(self, tenOne, tenTwo, intKernelSize):
+        _ensure_cupy()
+        if tenOne.is_cuda and cupy is not None:
             tenOut = tenOne.new_empty([tenOne.shape[0], intKernelSize ** 2, tenOne.shape[2], tenOne.shape[3]])
 
             cuda_launch(cuda_kernel('costvol_out', '''
@@ -302,6 +323,8 @@ class costvol_func(torch.autograd.Function):
                 args=[cuda_int32(tenOut.shape[0] * tenOut.shape[2] * tenOut.shape[3]), tenOne.data_ptr(), tenTwo.data_ptr(), intKernelSize, tenOut.data_ptr()],
                 stream=collections.namedtuple('Stream', 'ptr')(torch.cuda.current_stream().cuda_stream)
             )
+        else:
+            tenOut = _pytorch_costvol(tenOne, tenTwo, intKernelSize)
 
         self.save_for_backward(tenOne, tenTwo)
         self.intKernelSize = intKernelSize

gimm_vfi_arch/generalizable_INR/modules/softsplat.py

@@ -9,7 +9,10 @@
 # --------------------------------------------------------
 
 import collections
-import cupy
+try:
+    import cupy
+except ImportError:
+    cupy = None
 import os
 import re
 import torch
@@ -260,31 +263,78 @@ def cuda_kernel(strFunction: str, strKernel: str, objVariables: typing.Dict):
 # end
 
 
-@cupy.memoize(for_each_device=True)
+_cuda_launch_cache = {}
+
 @torch.compiler.disable()
 def cuda_launch(strKey: str):
-    try:
-        os.environ.setdefault("CUDA_HOME", cupy.cuda.get_cuda_path())
-    except Exception:
+    if strKey not in _cuda_launch_cache:
         if "CUDA_HOME" not in os.environ:
-            raise RuntimeError("'CUDA_HOME' not set, unable to find cuda-toolkit installation.")
-    
+            try:
+                cuda_path = cupy.cuda.get_cuda_path()
+            except Exception:
+                cuda_path = None
+            if cuda_path is None:
+                cuda_path = "/usr/local/cuda"
+            os.environ["CUDA_HOME"] = cuda_path
         strKernel = objCudacache[strKey]["strKernel"]
         strFunction = objCudacache[strKey]["strFunction"]
-    
-    return cupy.RawModule(
+        _cuda_launch_cache[strKey] = cupy.RawModule(
             code=strKernel,
             options=(
                 "-I " + os.environ["CUDA_HOME"],
                 "-I " + os.environ["CUDA_HOME"] + "/include",
             ),
         ).get_function(strFunction)
+    return _cuda_launch_cache[strKey]
 
 
 
 ##########################################################
 
 
+def _pytorch_softsplat(tenIn, tenFlow):
+    """Pure-PyTorch forward warp via bilinear splatting (scatter_add)."""
+    B, C, H, W = tenIn.shape
+    tenOut = tenIn.new_zeros(B, C, H, W)
+
+    grid_y, grid_x = torch.meshgrid(
+        torch.arange(H, device=tenIn.device, dtype=tenIn.dtype),
+        torch.arange(W, device=tenIn.device, dtype=tenIn.dtype),
+        indexing='ij',
+    )
+
+    flt_x = grid_x.unsqueeze(0) + tenFlow[:, 0, :, :]
+    flt_y = grid_y.unsqueeze(0) + tenFlow[:, 1, :, :]
+
+    valid = torch.isfinite(flt_x) & torch.isfinite(flt_y)
+    flt_x = torch.where(valid, flt_x, torch.zeros_like(flt_x))
+    flt_y = torch.where(valid, flt_y, torch.zeros_like(flt_y))
+
+    nw_x = flt_x.floor().long()
+    nw_y = flt_y.floor().long()
+    frac_x = flt_x - nw_x.to(flt_x.dtype)
+    frac_y = flt_y - nw_y.to(flt_y.dtype)
+
+    w_nw = (1.0 - frac_x) * (1.0 - frac_y) * valid
+    w_ne = frac_x * (1.0 - frac_y) * valid
+    w_sw = (1.0 - frac_x) * frac_y * valid
+    w_se = frac_x * frac_y * valid
+
+    out_flat = tenOut.view(B, C, -1)
+
+    for dx, dy, w in [(0, 0, w_nw), (1, 0, w_ne), (0, 1, w_sw), (1, 1, w_se)]:
+        tx = nw_x + dx
+        ty = nw_y + dy
+        in_bounds = (tx >= 0) & (tx < W) & (ty >= 0) & (ty < H)
+        w_masked = w * in_bounds
+        idx = (ty.clamp(0, H - 1) * W + tx.clamp(0, W - 1))
+        idx = idx.unsqueeze(1).expand_as(tenIn)
+        weighted = tenIn * w_masked.unsqueeze(1)
+        out_flat.scatter_add_(2, idx.reshape(B, C, -1), weighted.reshape(B, C, -1))
+
+    return tenOut
+
+
 @torch.compiler.disable()
 def softsplat(tenIn, tenFlow, tenMetric, strMode, return_norm=False):
     assert strMode.split("-")[0] in ["sum", "avg", "linear", "softmax"]
@@ -366,7 +416,7 @@ class softsplat_func(torch.autograd.Function):
             [tenIn.shape[0], tenIn.shape[1], tenIn.shape[2], tenIn.shape[3]]
         )
 
-        if tenIn.is_cuda == True:
+        if tenIn.is_cuda and cupy is not None:
             cuda_launch(
                 cuda_kernel(
                     "softsplat_out",
@@ -439,8 +489,8 @@ class softsplat_func(torch.autograd.Function):
                 ),
             )
 
-        elif tenIn.is_cuda != True:
-            assert False
+        else:
+            tenOut = _pytorch_softsplat(tenIn, tenFlow)
 
         # end
 

install.py

@@ -8,44 +8,39 @@ def get_cupy_package():
     try:
         import torch
         if not torch.cuda.is_available():
-            print("[Tween] WARNING: CUDA not available. cupy requires CUDA.")
             return None
         cuda_version = torch.version.cuda
         if cuda_version is None:
-            print("[Tween] WARNING: PyTorch has no CUDA version info.")
             return None
         major = int(cuda_version.split(".")[0])
         cupy_pkg = f"cupy-cuda{major}x"
-        print(f"[Tween] Detected CUDA {cuda_version}, will use {cupy_pkg}")
         return cupy_pkg
-    except Exception as e:
-        print(f"[Tween] WARNING: Could not detect CUDA version: {e}")
+    except Exception:
         return None
 
 
-def update_requirements(cupy_pkg):
-    """Write the correct cupy package into requirements.txt."""
-    requirements_path = os.path.join(os.path.dirname(__file__), "requirements.txt")
-    lines = []
-    if os.path.exists(requirements_path):
-        with open(requirements_path, "r") as f:
-            lines = [l.rstrip() for l in f if not l.strip().startswith("cupy")]
-    if cupy_pkg and cupy_pkg not in lines:
-        lines.append(cupy_pkg)
-    with open(requirements_path, "w") as f:
-        f.write("\n".join(lines) + "\n")
-
-
 def install():
-    cupy_pkg = get_cupy_package()
-    if cupy_pkg:
-        update_requirements(cupy_pkg)
-
+    # Install core requirements first
     requirements_path = os.path.join(os.path.dirname(__file__), "requirements.txt")
     subprocess.check_call([
         sys.executable, "-m", "pip", "install", "-r", requirements_path
     ])
 
+    # Try to install cupy for NVIDIA users (optional, improves performance)
+    cupy_pkg = get_cupy_package()
+    if cupy_pkg:
+        try:
+            subprocess.check_call([
+                sys.executable, "-m", "pip", "install", cupy_pkg
+            ])
+            print(f"[Tween] cupy installed ({cupy_pkg}) — fast CUDA kernels enabled")
+        except subprocess.CalledProcessError:
+            print(f"[Tween] WARNING: Could not install {cupy_pkg}. "
+                  f"BIM-VFI, SGM-VFI, and GIMM-VFI will use slower PyTorch fallback.")
+    else:
+        print("[Tween] cupy skipped (no NVIDIA CUDA). "
+              "BIM-VFI, SGM-VFI, and GIMM-VFI will use PyTorch fallback.")
+
 
 if __name__ == "__main__":
     install()

nodes.py

@@ -19,26 +19,6 @@ from .gimm_vfi_arch import clear_gimm_caches
 logger = logging.getLogger("Tween")
 
 
-def _check_cupy(model_name):
-    """Raise a clear error if cupy is not installed."""
-    try:
-        import cupy  # noqa: F401
-    except ImportError:
-        try:
-            cuda_ver = torch.version.cuda or "unknown"
-            major = int(cuda_ver.split(".")[0])
-            cupy_pkg = f"cupy-cuda{major}x"
-        except Exception:
-            cuda_ver = "unknown"
-            cupy_pkg = "cupy-cuda12x  # adjust to your CUDA version"
-        raise RuntimeError(
-            f"{model_name} requires cupy but it is not installed.\n\n"
-            f"Your PyTorch CUDA version: {cuda_ver}\n\n"
-            f"Install it with:\n"
-            f"  pip install {cupy_pkg}\n\n"
-            f"If you are unsure of your CUDA version, run:\n"
-            f"  python -c \"import torch; print(torch.version.cuda)\""
-        )
 
 
 def _get_system_ram_gb():
@@ -206,7 +186,6 @@ class LoadBIMVFIModel:
     CATEGORY = "video/BIM-VFI"
 
     def load_model(self, model_path, auto_pyr_level, pyr_level):
-        _check_cupy("BIM-VFI")
         full_path = os.path.join(MODEL_DIR, model_path)
 
         if not os.path.exists(full_path):
@@ -1374,7 +1353,6 @@ class LoadSGMVFIModel:
     CATEGORY = "video/SGM-VFI"
 
     def load_model(self, model_path, tta, num_key_points):
-        _check_cupy("SGM-VFI")
         full_path = os.path.join(SGM_MODEL_DIR, model_path)
 
         if not os.path.exists(full_path):
@@ -1801,7 +1779,6 @@ class LoadGIMMVFIModel:
     CATEGORY = "video/GIMM-VFI"
 
     def load_model(self, model_path, ds_factor):
-        _check_cupy("GIMM-VFI")
         full_path = os.path.join(GIMM_MODEL_DIR, model_path)
 
         # Auto-download main model if missing

sgm_vfi_arch/softsplat.py

@@ -1,7 +1,10 @@
 #!/usr/bin/env python
 
 import collections
-import cupy
+try:
+    import cupy
+except ImportError:
+    cupy = None
 import os
 import re
 import torch
@@ -216,20 +219,76 @@ def cuda_kernel(strFunction:str, strKernel:str, objVariables:typing.Dict):
 # end
 
 
-@cupy.memoize(for_each_device=True)
-def cuda_launch(strKey:str):
-    if 'CUDA_HOME' not in os.environ:
-        os.environ['CUDA_HOME'] = cupy.cuda.get_cuda_path()
-    # end
+_cuda_launch_cache = {}
 
-    return cupy.RawKernel(objCudacache[strKey]['strKernel'], objCudacache[strKey]['strFunction'],
-                          options=tuple(['-I ' + os.environ['CUDA_HOME'], '-I ' + os.environ['CUDA_HOME'] + '/include']))
+def cuda_launch(strKey:str):
+    if strKey not in _cuda_launch_cache:
+        if 'CUDA_HOME' not in os.environ:
+            try:
+                cuda_path = cupy.cuda.get_cuda_path()
+            except Exception:
+                cuda_path = None
+            if cuda_path is None:
+                cuda_path = '/usr/local/cuda'
+            os.environ['CUDA_HOME'] = cuda_path
+        _cuda_launch_cache[strKey] = cupy.RawKernel(
+            objCudacache[strKey]['strKernel'],
+            objCudacache[strKey]['strFunction'],
+            options=tuple(['-I ' + os.environ['CUDA_HOME'],
+                           '-I ' + os.environ['CUDA_HOME'] + '/include'])
+        )
+    return _cuda_launch_cache[strKey]
 # end
 
 
 ##########################################################
 
 
+def _pytorch_softsplat(tenIn, tenFlow):
+    """Pure-PyTorch forward warp via bilinear splatting (scatter_add)."""
+    B, C, H, W = tenIn.shape
+    tenOut = tenIn.new_zeros(B, C, H, W)
+
+    grid_y, grid_x = torch.meshgrid(
+        torch.arange(H, device=tenIn.device, dtype=tenIn.dtype),
+        torch.arange(W, device=tenIn.device, dtype=tenIn.dtype),
+        indexing='ij',
+    )
+
+    flt_x = grid_x.unsqueeze(0) + tenFlow[:, 0, :, :]
+    flt_y = grid_y.unsqueeze(0) + tenFlow[:, 1, :, :]
+
+    valid = torch.isfinite(flt_x) & torch.isfinite(flt_y)
+    flt_x = torch.where(valid, flt_x, torch.zeros_like(flt_x))
+    flt_y = torch.where(valid, flt_y, torch.zeros_like(flt_y))
+
+    nw_x = flt_x.floor().long()
+    nw_y = flt_y.floor().long()
+    frac_x = flt_x - nw_x.to(flt_x.dtype)
+    frac_y = flt_y - nw_y.to(flt_y.dtype)
+
+    w_nw = (1.0 - frac_x) * (1.0 - frac_y) * valid
+    w_ne = frac_x * (1.0 - frac_y) * valid
+    w_sw = (1.0 - frac_x) * frac_y * valid
+    w_se = frac_x * frac_y * valid
+
+    out_flat = tenOut.view(B, C, -1)
+    in_flat = tenIn
+
+    for dx, dy, w in [(0, 0, w_nw), (1, 0, w_ne), (0, 1, w_sw), (1, 1, w_se)]:
+        tx = nw_x + dx
+        ty = nw_y + dy
+        in_bounds = (tx >= 0) & (tx < W) & (ty >= 0) & (ty < H)
+        w_masked = w * in_bounds
+        idx = (ty.clamp(0, H - 1) * W + tx.clamp(0, W - 1))
+        idx = idx.unsqueeze(1).expand_as(in_flat)
+        weighted = in_flat * w_masked.unsqueeze(1)
+        out_flat.scatter_add_(2, idx.reshape(B, C, -1), weighted.reshape(B, C, -1))
+
+    return tenOut
+# end
+
+
 def softsplat(tenIn:torch.Tensor, tenFlow:torch.Tensor, tenMetric:torch.Tensor, strMode:str):
     assert(strMode.split('-')[0] in ['sum', 'avg', 'linear', 'soft'])
 
@@ -281,7 +340,7 @@ class softsplat_func(torch.autograd.Function):
     def forward(self, tenIn, tenFlow):
         tenOut = tenIn.new_zeros([tenIn.shape[0], tenIn.shape[1], tenIn.shape[2], tenIn.shape[3]])
 
-        if tenIn.is_cuda == True:
+        if tenIn.is_cuda and cupy is not None:
             cuda_launch(cuda_kernel('softsplat_out', '''
                 extern "C" __global__ void __launch_bounds__(512) softsplat_out(
                     const int n,
@@ -345,8 +404,8 @@ class softsplat_func(torch.autograd.Function):
                 stream=collections.namedtuple('Stream', 'ptr')(torch.cuda.current_stream().cuda_stream)
             )
 
-        elif tenIn.is_cuda != True:
-            assert(False)
+        else:
+            tenOut = _pytorch_softsplat(tenIn, tenFlow)
 
         # end