feat(hunyuan15): add single-GPU CUDA text-to-video

DEVELOPER.md

@@ -100,7 +100,8 @@ difflet/
 └── models/          flux/ wan/ hunyuan_video/ qwen_image/ ltx_2/ — each:
                      modeling, pipeline, application, entry, checkpoint
                      (where needed), and spec.py for models on the GPU flow. HunyuanVideo 1.5 is served from
-                     hunyuan_video/ (model_version="1.5").
+                     hunyuan_video/ (model_version="1.5"); on CUDA it has its own path,
+                     hunyuan_video_15/spec.py and backends/cuda/hunyuan_video_15/.
 ```
 
 `difflet/backends/trainium/{core,modules}` follows upstream Neuron coding style so periodic
@@ -251,7 +252,8 @@ mesh `{dp, cfg, cp, tp}` (tp innermost, dp outermost):
   true-CFG models (Wan, LTX-2, HunyuanVideo 1.5) — guidance-distilled models (Flux,
   HunyuanVideo) and Qwen-Image run a single forward pass and reject it. HunyuanVideo 1.5 is
   not guidance-distilled: its guider runs classifier-free guidance at 6.0 on every step, and
-  `--cfg-parallel` (TP2 × CFG2) is its default layout.
+  `--cfg-parallel` (TP2 × CFG2) is its default layout on Trainium. On CUDA it runs both CFG
+  branches in sequence on one GPU, and `--cfg-parallel` is refused.
 - `sp_enabled` — Megatron-style sequence parallelism: shards the otherwise-replicated
   norm/modulation/residual regions along the sequence axis across the *existing*
   tensor-parallel group (reduce-scatter replaces the row-parallel all-reduce). Adds no mesh

NOTICE

@@ -52,7 +52,7 @@ This product includes software derived from the following third-party projects:
   `scripts/rebase_from_nxdi.py` for the rebase workflow.
 
 ================================================================================
-2. HuggingFace diffusers — pipeline extensions and LTX-2 attention
+2. HuggingFace diffusers — pipeline extensions, LTX-2 attention and HunyuanVideo 1.5
 ================================================================================
 
   Source:  https://github.com/huggingface/diffusers
@@ -66,6 +66,12 @@ This product includes software derived from the following third-party projects:
   LTX2VideoTransformer3DModel attention processor under Apache License 2.0.
   Copyright 2025 The Lightricks team and The HuggingFace Team.
   The file retains the upstream license header and identifies the adaptation.
+  The HunyuanVideo 1.5 CUDA transformer,
+  difflet/backends/cuda/hunyuan_video_15/modeling.py, adapts portions of
+  diffusers' models/transformers/transformer_hunyuan_video15.py. That file keeps
+  the upstream copyright (2025 The Hunyuan Team and The HuggingFace Team) and its
+  Apache License 2.0 header; Difflet replaces its projections and attention and
+  keeps the checkpoint layout.
 
 ================================================================================
 3. xDiT (xfuser) — design inspiration only, no code

README.md

@@ -124,6 +124,7 @@ see the per-model limits below.
 | Wan 2.2 | ✅ video | ✅ video | ✅ | ❌ | ⚠️ ⁵ | ✅ ⁶ | ❌ | ❌ |
 | HunyuanVideo (original) | ⚠️ experimental ⁷ | ⚠️ experimental ⁷ | ⚠️ execution observed ⁷ | ❌ | Not measured | — ¹ | ❌ | ❌ |
 | LTX-2 | ⚠️ video, CPU checks ⁸ | ⚠️ video, CPU checks ⁸ | ⚠️ Python qualified ⁸ | ❌ | Not measured | ❌ | ❌ | ❌ |
+| HunyuanVideo 1.5 | ⚠️ experimental ⁹ | ⚠️ experimental ⁹ | ✅ ⁹ | ❌ | Not measured | ❌ | ❌ | ❌ |
 
 An option the GPU backend does not support is refused with a message naming it, not ignored.
 TeaCache (adaptive), TAEF1, and `--host-vae` are refused on GPUs for these models.
@@ -148,6 +149,12 @@ TeaCache (adaptive), TAEF1, and `--host-vae` are refused on GPUs for these model
    and performance checks on B200. CLI/HTTP video-only paths have CPU coverage; GPU frontend
    validation is not claimed. The opt-in profile requires explicit TP=1 and a pre-CUDA startup
    setting; see the [LTX-2 CUDA guide](docs/verification/ltx2-cuda.md).
+9. Experimental single-GPU BF16 text-to-video for
+   `hunyuanvideo-community/HunyuanVideo-1.5-Diffusers-480p_t2v`, separate from the Trainium
+   HunyuanVideo 1.5 path. Classifier-free guidance runs both branches in sequence on one GPU.
+   On one RTX PRO 6000, eager and compiled output meet a relative accuracy bar against
+   vLLM-Omni and warm latency is within 10% of it; see the
+   [HunyuanVideo 1.5 CUDA guide](docs/verification/hunyuanvideo15-cuda.md).
 
 Measured timings against stock diffusers, and the full pass/limit matrix, are in
 [docs/verification/gpu-backend-results.md](docs/verification/gpu-backend-results.md).
@@ -313,9 +320,9 @@ LTX-2 also has an experimental single-GPU CUDA path with joint audiovisual Pytho
 and an opt-in native-compatible BF16 profile. See the [LTX-2 CUDA guide](docs/verification/ltx2-cuda.md)
 for its separate runtime, explicit TP=1/startup requirements and qualified eager/compiled results.
 
-For HunyuanVideo CUDA, pass `parallel=DiffletParallelConfig(tp_degree=1)` explicitly.
-Omitting `parallel` preserves the model-specific registry default, including HunyuanVideo's
-Trainium topology. GPU Python API paths have unit coverage; see the
+For HunyuanVideo and HunyuanVideo 1.5 on CUDA, pass `parallel=DiffletParallelConfig(tp_degree=1)`
+explicitly. Omitting `parallel` preserves the model-specific registry default, including the
+HunyuanVideo models' Trainium topologies. GPU Python API paths have unit coverage; see the
 [per-model results](docs/verification/gpu-backend-results.md) for hardware coverage.
 
 ## Choose a topology

difflet/backends/cuda/hunyuan_video_15/__init__.py

@@ -0,0 +1 @@
+"""Standard HunyuanVideo 1.5 on CUDA."""

difflet/backends/cuda/hunyuan_video_15/application.py

@@ -0,0 +1,175 @@
+"""Standard 1.5 conditioning, CFG and scheduler semantics from the Diffusers driver."""
+
+from __future__ import annotations
+
+from contextlib import contextmanager
+
+import torch
+from diffusers.models.modeling_outputs import Transformer2DModelOutput
+
+from difflet.backends.cuda.hunyuan_video_15.conditioning import configure_conditioning
+from difflet.backends.cuda.hunyuan_video_15.dit_config import HunyuanVideo15DiTConfig
+from difflet.backends.gpu.application import GpuApplication
+from difflet.models.hunyuan_video_15.spec import SPEC
+
+
+class HunyuanVideo15DiTAdapter(torch.nn.Module):
+    def __init__(self, component):
+        super().__init__()
+        self.component = component
+        self.config = component.dit_config
+
+    @property
+    def dtype(self):
+        return self.component.dtype
+
+    @property
+    def device(self):
+        return self.component.device
+
+    @contextmanager
+    def cache_context(self, name):
+        yield
+
+    def forward(
+        self,
+        hidden_states,
+        timestep,
+        encoder_hidden_states,
+        encoder_attention_mask,
+        encoder_hidden_states_2,
+        encoder_attention_mask_2,
+        image_embeds,
+        attention_kwargs=None,
+        return_dict=True,
+    ):
+        # Every input is cast to the transformer's dtype, as Diffusers' HunyuanVideo 1.5 driver
+        # does. Unlike the original HunyuanVideo adapter, the timestep is not kept in FP32.
+        values = (
+            hidden_states,
+            timestep,
+            encoder_hidden_states,
+            encoder_attention_mask,
+            encoder_hidden_states_2,
+            encoder_attention_mask_2,
+            image_embeds,
+        )
+        inputs = tuple(
+            t.to(device=self.device, dtype=torch.bool if i in (3, 5) else self.dtype)
+            for i, t in enumerate(values)
+        )
+        result = self.component(*inputs, return_dict=False)[0]
+        return Transformer2DModelOutput(sample=result) if return_dict else (result,)
+
+
+class CudaHunyuanVideo15Application(GpuApplication):
+    spec = SPEC
+
+    def _configure(self, kwargs):
+        self.local_files_only = bool(kwargs.get("local_files_only", True))
+        self.pipe = None
+        self.adapter = None
+        self._guider_config = None
+
+    def _build_transformers(self):
+        from difflet.backends.cuda.hunyuan_video_15.transformer import (
+            CudaHunyuanVideo15TransformerApplication,
+        )
+
+        self.config = HunyuanVideo15DiTConfig.from_pretrained(
+            self.transformer_path, **self.shape, torch_dtype=self.dtype
+        )
+        self.transformer = CudaHunyuanVideo15TransformerApplication(
+            model_path=self.transformer_path,
+            config=self.config,
+            device=self.device,
+            torch_compile=self.torch_compile,
+        )
+        self.adapter = HunyuanVideo15DiTAdapter(self.transformer)
+
+    def load_companions(self):
+        from diffusers import HunyuanVideo15Pipeline
+
+        base = HunyuanVideo15Pipeline.from_pretrained(
+            self.model_path,
+            transformer=None,
+            torch_dtype=self.dtype,
+            local_files_only=self.local_files_only,
+        )
+        # Diffusers' native overlapping spatial tiles (256 px, 25% overlap) keep the decode
+        # within one GPU at 121 frames.
+        base.vae.enable_tiling()
+        for name in ("text_encoder", "text_encoder_2", "vae"):
+            getattr(base, name).to(self.device).requires_grad_(False).eval()
+        self.pipe = HunyuanVideo15Pipeline(**{**base.components, "transformer": self.adapter})
+        configure_conditioning(self.pipe)
+        self._guider_config = dict(base.guider.config)
+
+    def dit_input_contract(self):
+        c = self._require_transformer("input contract").dit_config
+        shapes = (
+            ("hidden_states", (1, c.in_channels, c.latent_frames, c.latent_height, c.latent_width)),
+            ("timestep", (1,)),
+            ("encoder_hidden_states", (1, c.text_seq_len, c.text_embed_dim)),
+            ("encoder_attention_mask", (1, c.text_seq_len)),
+            ("encoder_hidden_states_2", (1, c.text_2_seq_len, c.text_embed_2_dim)),
+            ("encoder_attention_mask_2", (1, c.text_2_seq_len)),
+            ("image_embeds", (1, c.image_tokens, c.image_embed_dim)),
+        )
+        return {
+            name: {"shape": shape, "dtype": torch.bool if "mask" in name else self.dtype}
+            for name, shape in shapes
+        }
+
+    def generate(
+        self,
+        prompt,
+        *,
+        negative_prompt=None,
+        num_inference_steps=50,
+        guidance_scale=6.0,
+        seed=None,
+        latents=None,
+        output_type="pt",
+        height=None,
+        width=None,
+        num_frames=None,
+    ):
+        from diffusers import ClassifierFreeGuidance
+
+        if self.pipe is None:
+            raise RuntimeError("application is not loaded; call load() first")
+        if not isinstance(prompt, str):
+            raise ValueError("HunyuanVideo 1.5 generates one prompt per request")
+        if num_inference_steps < 1:
+            raise ValueError("num_inference_steps must be at least 1")
+        for key, value in (("height", height), ("width", width), ("num_frames", num_frames)):
+            if value is not None and value != self.shape[key]:
+                raise ValueError(f"{key} differs from the loaded shape {self.shape[key]}")
+        if output_type not in ("pt", "latent"):
+            raise ValueError("HunyuanVideo 1.5 output_type must be pt or latent")
+        generator = None
+        if latents is not None:
+            c = self.config
+            expected = (1, c.out_channels, c.latent_frames, c.latent_height, c.latent_width)
+            if tuple(latents.shape) != expected:
+                raise ValueError(f"latents must have shape {expected}")
+        elif seed is not None:
+            generator = torch.Generator(device=self.device).manual_seed(int(seed))
+        self.pipe.guider = ClassifierFreeGuidance.from_config(
+            {**self._guider_config, "guidance_scale": float(guidance_scale)}
+        )
+        output = self.pipe(
+            prompt=prompt,
+            negative_prompt=negative_prompt,
+            height=self.shape["height"],
+            width=self.shape["width"],
+            num_frames=self.shape["num_frames"],
+            num_inference_steps=num_inference_steps,
+            generator=generator,
+            latents=latents,
+            output_type=output_type,
+        ).frames
+        if output_type == "latent":
+            return output
+        return output.permute(0, 2, 1, 3, 4).to(device="cpu", dtype=torch.float32) * 2 - 1

difflet/backends/cuda/hunyuan_video_15/conditioning.py

@@ -0,0 +1,42 @@
+"""Prompt conditioning as the reference computes it: device-local Qwen rotary
+frequencies, and cuDNN SDPA off while prompts are encoded."""
+
+from functools import wraps
+
+import torch
+
+
+def initialize_qwen_rotary(qwen):
+    """Recompute the nonpersistent FP32 rotary frequencies on the encoder's device.
+
+    Frequencies computed on the CPU and moved to CUDA differ in the last bits from the
+    device's own FP32 power arithmetic, which the reference conditioning uses.
+    """
+    rotary = qwen.rotary_emb
+    if rotary.config.rope_parameters.get("rope_type", "default") != "default":
+        raise NotImplementedError("HunyuanVideo 1.5 requires default Qwen rotary embeddings")
+    with torch.no_grad(), torch.device(next(qwen.parameters()).device):
+        inv_freq, _ = rotary.compute_default_rope_parameters(rotary.config)
+    for name in ("inv_freq", "original_inv_freq"):
+        setattr(rotary, name, inv_freq.clone())
+
+
+def configure_conditioning(pipe):
+    """Encode prompts with cuDNN SDPA disabled, as the reference conditioning does.
+
+    Only ``encode_prompt`` changes the policy, and it restores the caller's afterwards; the
+    denoiser keeps its own. No model class or learned weight changes.
+    """
+    initialize_qwen_rotary(pipe.text_encoder)
+    original = pipe.encode_prompt
+
+    @wraps(original)
+    def encode_prompt(*args, **kwargs):
+        previous = torch.backends.cuda.cudnn_sdp_enabled()
+        try:
+            torch.backends.cuda.enable_cudnn_sdp(False)
+            return original(*args, **kwargs)
+        finally:
+            torch.backends.cuda.enable_cudnn_sdp(previous)
+
+    pipe.encode_prompt = encode_prompt

difflet/backends/cuda/hunyuan_video_15/dit_config.py

@@ -0,0 +1,78 @@
+"""The HunyuanVideo 1.5 text-to-video transformer config and its resident shape."""
+
+from __future__ import annotations
+
+import json
+from dataclasses import dataclass, fields
+from pathlib import Path
+
+import torch
+
+
+@dataclass
+class HunyuanVideo15DiTConfig:
+    in_channels: int = 65
+    out_channels: int = 32
+    num_attention_heads: int = 16
+    attention_head_dim: int = 128
+    num_layers: int = 54
+    num_refiner_layers: int = 2
+    mlp_ratio: float = 4.0
+    patch_size: int = 1
+    patch_size_t: int = 1
+    qk_norm: str = "rms_norm"
+    text_embed_dim: int = 3584
+    text_embed_2_dim: int = 1472
+    image_embed_dim: int = 1152
+    rope_theta: float = 256.0
+    rope_axes_dim: tuple[int, ...] = (16, 56, 56)
+    target_size: int = 640
+    task_type: str = "t2v"
+    use_meanflow: bool = False
+    height: int = 480
+    width: int = 848
+    num_frames: int = 121
+    text_seq_len: int = 1000
+    text_2_seq_len: int = 256
+    image_tokens: int = 729
+    torch_dtype: torch.dtype = torch.bfloat16
+
+    def __post_init__(self):
+        # The one owner of the resident shape: the application builds this config before any
+        # weights load.
+        if min(self.height, self.width, self.num_frames) <= 0:
+            raise ValueError("HunyuanVideo 1.5 height, width and num_frames must be positive")
+        if self.height % 16 or self.width % 16:
+            raise ValueError("HunyuanVideo 1.5 height and width must be multiples of 16")
+        if (self.num_frames - 1) % 4:
+            raise ValueError("HunyuanVideo 1.5 num_frames must be 4n+1")
+        if self.task_type != "t2v" or self.use_meanflow:
+            raise NotImplementedError("HunyuanVideo 1.5 supports standard text-to-video only")
+        if self.patch_size != 1 or self.patch_size_t != 1 or self.qk_norm != "rms_norm":
+            raise NotImplementedError(
+                "HunyuanVideo 1.5 supports patch size 1 with RMSNorm Q/K normalization only"
+            )
+        self.rope_axes_dim = tuple(self.rope_axes_dim)
+
+    @property
+    def latent_height(self):
+        return self.height // 16
+
+    @property
+    def latent_width(self):
+        return self.width // 16
+
+    @property
+    def latent_frames(self):
+        return (self.num_frames - 1) // 4 + 1
+
+    @classmethod
+    def from_pretrained(cls, transformer_dir, **overrides):
+        body = json.loads((Path(transformer_dir) / "config.json").read_text())
+        if (
+            body.get("_class_name", "HunyuanVideo15Transformer3DModel")
+            != "HunyuanVideo15Transformer3DModel"
+        ):
+            raise NotImplementedError("Unsupported HunyuanVideo 1.5 transformer class")
+        known = {f.name for f in fields(cls)}
+        return cls(**{**{k: v for k, v in body.items() if k in known}, **overrides})