From 0d9b13bf0a8b1590c60799a2f8a6b9ac673f9e5a Mon Sep 17 00:00:00 2001
From: Alan Liu <zaoxingliu@gmail.com>
Date: Mon, 24 Aug 2026 22:27:41 -0400
Subject: [PATCH 1/5] Implement project updates

---
 benchmarks/bench_clickhouse_host.py     | 352 ++++++++++++++++++++++--
 docs/benchmarks.md                      |  58 +++-
 docs/clickhouse-offload-pipeline.html   | 153 ++++++++++
 docs/config.md                          |  19 +-
 docs/integration-api-v1.md              |  39 ++-
 native/csrc/bindings.cpp                |  50 +++-
 native/csrc/clickhouse_client.cpp       | 110 +++++++-
 native/csrc/clickhouse_client.h         |  58 ++++
 native/csrc/dmx_host_engine.h           |  69 ++++-
 native/csrc/ring/drain_thread.cpp       | 166 ++++++++---
 native/csrc/ring/drain_thread.h         |  29 +-
 native/csrc/ring/p2p_thread.cpp         |  76 ++---
 native/csrc/ring/ring_config.h          |   5 +-
 native/csrc/ring/ring_engine.cu         |  39 ++-
 native/csrc/ring/ring_engine.h          |   2 +
 native/csrc/ring/ring_engine_py.cu      |  80 ++----
 native/csrc/ring/ring_engine_py.h       |  15 +-
 native/csrc/ring/task_entry.h           |   2 +-
 native/csrc/ring/task_ring.cuh          |   6 +-
 src/dmi/engine.py                       |  42 ++-
 src/dmi/hooks/point.py                  |   8 +-
 src/dmi/transport/ring.py               |   4 +-
 tests/native/ring/Makefile              |   3 +-
 tests/native/ring/test_producer.cu      |  27 ++
 tests/native/ring/test_ring_engine.cu   | 141 +++++++++-
 tests/test_clickhouse_host_benchmark.py | 199 +++++++++++++-
 tests/test_cpu_native_build.py          |  43 +++
 tests/test_engine_runtime_api.py        |  97 ++++++-
 28 files changed, 1610 insertions(+), 282 deletions(-)
 create mode 100644 docs/clickhouse-offload-pipeline.html

diff --git a/benchmarks/bench_clickhouse_host.py b/benchmarks/bench_clickhouse_host.py
index 01aae389e..4cee98a74 100644
--- a/benchmarks/bench_clickhouse_host.py
+++ b/benchmarks/bench_clickhouse_host.py
@@ -6,12 +6,15 @@
 import json
 import math
 import os
+import random
 import re
 import resource
+import statistics
 import sys
+import threading
 import time
 import uuid
-from dataclasses import asdict, dataclass, field
+from dataclasses import asdict, dataclass, field, replace
 from datetime import datetime
 from typing import Any, Callable, Sequence
 
@@ -74,6 +77,8 @@ class BenchmarkConfig:
     compression: str = "lz4"
     async_insert: bool = False
     drain_timeout_seconds: float = 300.0
+    socket_timeout_seconds: float = 30.0
+    server_sample_interval_ms: int = 50
     host: str = "localhost"
     port: int = 9000
     user: str = "default"
@@ -109,6 +114,10 @@ def __post_init__(self) -> None:
             raise ValueError("max_linger_ms must be a finite non-negative number")
         if not math.isfinite(self.drain_timeout_seconds) or self.drain_timeout_seconds <= 0:
             raise ValueError("drain_timeout_seconds must be a finite positive number")
+        if not math.isfinite(self.socket_timeout_seconds) or self.socket_timeout_seconds <= 0:
+            raise ValueError("socket_timeout_seconds must be a finite positive number")
+        if self.server_sample_interval_ms < 0:
+            raise ValueError("server_sample_interval_ms must be non-negative")
         if not 1 <= self.port <= 65535:
             raise ValueError("port must be between 1 and 65535")
         if self.dtype not in _DTYPE_BYTES:
@@ -155,8 +164,9 @@ class TrialMeasurement:
     total_seconds: float
     enqueue_latencies_ns: tuple[int, ...]
     process_cpu_seconds: float = 0.0
-    peak_rss_before_bytes: int = 0
-    peak_rss_after_bytes: int = 0
+    process_lifetime_peak_rss_bytes: int = 0
+    startup_seconds: float = 0.0
+    client_metrics: dict[str, Any] | None = None
 
     def as_dict(self) -> dict[str, Any]:
         logical_bytes = self.rows * self.payload_bytes
@@ -177,26 +187,163 @@ def rates(seconds: float) -> dict[str, float]:
         return {
             "rows": self.rows,
             "logical_payload_bytes": logical_bytes,
+            "startup_seconds": self.startup_seconds,
             "enqueue": enqueue,
             "drain_seconds": max(0.0, self.total_seconds - self.enqueue_seconds),
             "total": rates(self.total_seconds),
             "host_process": {
                 "cpu_seconds": self.process_cpu_seconds,
                 "effective_cpu_cores": self.process_cpu_seconds / self.total_seconds,
-                "peak_rss_bytes": self.peak_rss_after_bytes,
-                "peak_rss_growth_bytes": max(
-                    0, self.peak_rss_after_bytes - self.peak_rss_before_bytes
-                ),
+                "process_lifetime_peak_rss_bytes": self.process_lifetime_peak_rss_bytes,
             },
+            "client": self.client_metrics,
         }
 
 
 def build_client_settings(async_insert: bool) -> dict[str, int]:
     if not async_insert:
-        return {}
+        return {"async_insert": 0}
     return {"async_insert": 1, "wait_for_async_insert": 1}
 
 
+def parse_parallelism_sweep(value: str) -> tuple[int, ...]:
+    try:
+        parsed = [int(part.strip()) for part in value.split(",") if part.strip()]
+    except ValueError as exc:
+        raise ValueError("parallelism values must be integers") from exc
+    if not parsed or any(value <= 0 for value in parsed):
+        raise ValueError("parallelism values must be positive")
+    return tuple(dict.fromkeys(parsed))
+
+
+def _client_metrics_as_dict(metrics: Any) -> dict[str, Any]:
+    workers = [
+        {
+            "worker_index": int(worker.worker_index),
+            "batches": int(worker.batches),
+            "rows": int(worker.rows),
+            "logical_bytes": int(worker.logical_bytes),
+            "insert_seconds": float(worker.insert_seconds),
+        }
+        for worker in metrics.workers
+    ]
+    return {
+        "expected_workers": int(metrics.expected_workers),
+        "ready_workers": int(metrics.ready_workers),
+        "active_inserts": int(metrics.active_inserts),
+        "peak_active_inserts": int(metrics.peak_active_inserts),
+        "batches": int(metrics.batches),
+        "rows": int(metrics.rows),
+        "logical_bytes": int(metrics.logical_bytes),
+        "insert_seconds": float(metrics.insert_seconds),
+        "workers": workers,
+    }
+
+
+class ServerTelemetrySampler:
+    _PROCESS_QUERY = "SELECT count() FROM system.processes WHERE query_kind = 'Insert'"
+    _METRICS_QUERY = """
+        SELECT metric, value
+        FROM system.metrics
+        WHERE metric IN ('Query', 'Merge', 'BackgroundMergesAndMutationsPoolTask',
+                         'TCPConnection')
+    """
+    _ASYNC_METRICS_QUERY = """
+        SELECT metric,
+               if(isFinite(value), value, arraySum(mapValues(key_values))) AS scalar_value
+        FROM system.asynchronous_metrics
+        WHERE metric IN ('OSUserTimeNormalized', 'OSSystemTimeNormalized',
+                         'OSIOWaitTimeNormalized', 'MemoryResident', 'LoadAverage1',
+                         'MaxPartCountForPartition', 'LongestRunningMerge',
+                         'BlockReadBytes', 'BlockWriteBytes',
+                         'NetworkReceiveBytes', 'NetworkSendBytes')
+    """
+
+    def __init__(self, client_factory: Callable[[], Any], interval_ms: int = 50):
+        if interval_ms <= 0:
+            raise ValueError("interval_ms must be positive")
+        self._client_factory = client_factory
+        self._interval_seconds = interval_ms / 1000.0
+        self._stop = threading.Event()
+        self._thread: threading.Thread | None = None
+        self._client: Any = None
+        self._lock = threading.Lock()
+        self._samples = 0
+        self._peak_active_inserts = 0
+        self._metric_values: dict[str, list[float]] = {}
+        self._errors: list[str] = []
+
+    def start(self) -> None:
+        if self._thread is not None:
+            return
+        self.sample_once()
+        if self._stop.is_set():
+            return
+        self._thread = threading.Thread(
+            target=self._run,
+            name="dmi-clickhouse-telemetry",
+            daemon=True,
+        )
+        self._thread.start()
+
+    def stop(self) -> None:
+        self._stop.set()
+        if self._thread is not None:
+            self._thread.join()
+        if not self._errors:
+            self.sample_once()
+        if self._client is not None:
+            try:
+                self._client.disconnect()
+            except Exception:
+                pass
+
+    def _run(self) -> None:
+        while not self._stop.is_set():
+            self.sample_once()
+            self._stop.wait(self._interval_seconds)
+
+    def sample_once(self) -> None:
+        try:
+            if self._client is None:
+                self._client = self._client_factory()
+            active = int(self._client.execute(self._PROCESS_QUERY)[0][0])
+            metrics = self._client.execute(self._METRICS_QUERY)
+            async_metrics = self._client.execute(self._ASYNC_METRICS_QUERY)
+            with self._lock:
+                self._samples += 1
+                self._peak_active_inserts = max(self._peak_active_inserts, active)
+                for name, value in metrics:
+                    self._metric_values.setdefault(str(name), []).append(float(value))
+                for name, value in async_metrics:
+                    key = f"async.{name}"
+                    self._metric_values.setdefault(key, []).append(float(value))
+        except Exception as exc:
+            with self._lock:
+                if not self._errors:
+                    self._errors.append(str(exc))
+            self._stop.set()
+
+    def snapshot(self) -> dict[str, Any]:
+        with self._lock:
+            return {
+                "samples": self._samples,
+                "peak_active_inserts": self._peak_active_inserts,
+                "metric_peaks": {
+                    name: max(values) for name, values in self._metric_values.items()
+                },
+                "metric_means": {
+                    name: statistics.fmean(values)
+                    for name, values in self._metric_values.items()
+                },
+                "metric_deltas": {
+                    name: values[-1] - values[0]
+                    for name, values in self._metric_values.items()
+                },
+                "errors": list(self._errors),
+            }
+
+
 def generate_payload_pool(config: BenchmarkConfig) -> list[Any]:
     import torch
 
@@ -229,6 +376,10 @@ def configure_stage(stage: Any, config: BenchmarkConfig) -> None:
     queue.high_watermark_items = config.queue_capacity_items
     queue.high_watermark_size = config.queue_capacity_bytes
     stage.ingress_policy.block = True
+    stage.ingress_policy.timeout_s = min(
+        config.socket_timeout_seconds,
+        config.drain_timeout_seconds,
+    )
 
 
 def _peak_rss_bytes() -> int:
@@ -256,6 +407,10 @@ def _build_engine(config: BenchmarkConfig) -> Any:
     clickhouse.client_side_compress = config.compression
     clickhouse.client_settings = build_client_settings(config.async_insert)
     clickhouse.index_granularity = config.index_granularity
+    timeout_ms = round(config.socket_timeout_seconds * 1000)
+    clickhouse.connect_timeout_ms = min(5000, timeout_ms)
+    clickhouse.send_timeout_ms = timeout_ms
+    clickhouse.receive_timeout_ms = timeout_ms
 
     stage = StageConfig.clickhouse_insert(
         clickhouse,
@@ -274,15 +429,29 @@ def _submit_and_drain(
     timeout_seconds: float,
     max_latency_samples: int = 10_000,
     clock: Callable[[], int] = time.perf_counter_ns,
+    deadline_clock: Callable[[], float] = time.monotonic,
+    sampler: ServerTelemetrySampler | None = None,
 ) -> TrialMeasurement:
-    engine.start()
-    cpu_start = time.process_time()
-    peak_rss_before = _peak_rss_bytes()
-    total_start = clock()
-    latencies = []
-    sample_stride = max(1, math.ceil(rows / max_latency_samples))
+    startup_start = clock()
+    sampler_started = False
+    deadline = deadline_clock() + timeout_seconds
     try:
+        engine.start()
+        if not engine.wait_until_ready(timeout_seconds):
+            engine.raise_if_failed()
+            raise TimeoutError(f"ClickHouse workers were not ready after {timeout_seconds:.1f}s")
+        startup_end = clock()
+        deadline = deadline_clock() + timeout_seconds
+        cpu_start = time.process_time()
+        if sampler is not None:
+            sampler.start()
+            sampler_started = True
+        total_start = clock()
+        latencies = []
+        sample_stride = max(1, math.ceil(rows / max_latency_samples))
         for row in range(rows):
+            if deadline_clock() >= deadline:
+                raise TimeoutError(f"ClickHouse benchmark exceeded {timeout_seconds:.1f}s")
             started = clock() if row % sample_stride == 0 else None
             engine.submit_direct(
                 model_id,
@@ -298,16 +467,28 @@ def _submit_and_drain(
                 latencies.append(clock() - started)
         enqueue_end = clock()
         engine.close_input()
-        if not engine.join(timeout_seconds):
+        remaining = max(0.0, deadline - deadline_clock())
+        if not engine.join(remaining):
             raise TimeoutError(f"ClickHouse drain exceeded {timeout_seconds:.1f}s")
         engine.raise_if_failed()
         total_end = clock()
+        if sampler is not None:
+            sampler.stop()
+            sampler_started = False
         cpu_end = time.process_time()
-        peak_rss_after = _peak_rss_bytes()
+        process_lifetime_peak_rss = _peak_rss_bytes()
+        client_metrics = _client_metrics_as_dict(engine.clickhouse_metrics())
     except BaseException:
         engine.request_abort()
-        engine.join(timeout_seconds)
+        remaining = max(0.0, deadline - deadline_clock())
+        try:
+            engine.join(remaining)
+        except BaseException:
+            pass
         raise
+    finally:
+        if sampler_started:
+            sampler.stop()
 
     return TrialMeasurement(
         rows=rows,
@@ -316,23 +497,27 @@ def _submit_and_drain(
         total_seconds=(total_end - total_start) / 1_000_000_000,
         enqueue_latencies_ns=tuple(latencies),
         process_cpu_seconds=cpu_end - cpu_start,
-        peak_rss_before_bytes=peak_rss_before,
-        peak_rss_after_bytes=peak_rss_after,
+        process_lifetime_peak_rss_bytes=process_lifetime_peak_rss,
+        startup_seconds=(startup_end - startup_start) / 1_000_000_000,
+        client_metrics=client_metrics,
     )
 
 
-def _connect(config: BenchmarkConfig) -> Any:
+def _connect(config: BenchmarkConfig, timeout_seconds: float | None = None) -> Any:
     try:
         from clickhouse_driver import Client
     except ImportError as exc:
         raise RuntimeError("clickhouse-driver is required for benchmark verification") from exc
 
+    socket_timeout = timeout_seconds or config.socket_timeout_seconds
     connection = {
         "host": config.host,
         "port": config.port,
         "user": config.user,
         "password": config.password,
         "secure": config.secure,
+        "connect_timeout": min(5.0, socket_timeout),
+        "send_receive_timeout": socket_timeout,
     }
     if config.create_database:
         bootstrap = Client(database="default", **connection)
@@ -441,13 +626,7 @@ def _safe_config(config: BenchmarkConfig) -> dict[str, Any]:
 
 
 def _ensure_table(client: Any, config: BenchmarkConfig) -> None:
-    """Explicitly create the capture table if it does not yet exist.
-
-    The native backend guards its own DDL with a process-wide once_flag, so a
-    second ``run()`` call in the same process will not recreate a table that
-    was dropped by a previous run.  Pre-creating the table here means every
-    ``run()`` starts with a live table regardless of the once_flag state.
-    """
+    """Create the trial table before native workers enter the timed path."""
     table = f"{quote_identifier(config.database)}.{quote_identifier(config.table)}"
     cols = ", ".join([
         f"{quote_identifier('model_id')} String",
@@ -497,6 +676,12 @@ def run(config: BenchmarkConfig) -> dict[str, Any]:
             client.execute(f"TRUNCATE TABLE {table}")
 
         started_at = _server_time(client)
+        sampler = None
+        if config.server_sample_interval_ms:
+            sampler = ServerTelemetrySampler(
+                lambda: _connect(config, min(2.0, config.socket_timeout_seconds)),
+                interval_ms=config.server_sample_interval_ms,
+            )
         measurement = _submit_and_drain(
             _build_engine(config),
             payloads,
@@ -504,7 +689,11 @@ def run(config: BenchmarkConfig) -> dict[str, Any]:
             measured_model_id,
             config.drain_timeout_seconds,
             config.latency_samples,
+            sampler=sampler,
         )
+        server_telemetry = sampler.snapshot() if sampler is not None else None
+        if server_telemetry and server_telemetry["errors"]:
+            warnings.append(f"server telemetry unavailable: {server_telemetry['errors'][0]}")
         verification = _verification(client, config, measured_model_id)
         parts = _parts_metrics(client, config)
         query_log, warning = _query_log_metrics(client, config, started_at)
@@ -514,10 +703,12 @@ def run(config: BenchmarkConfig) -> dict[str, Any]:
             "benchmark": "dmi_clickhouse_host",
             "server_version": server_version,
             "config": _safe_config(config),
+            "effective_client_settings": build_client_settings(config.async_insert),
             "measurement": measurement.as_dict(),
             "verification": verification,
             "parts": parts,
             "query_log": query_log,
+            "server_telemetry": server_telemetry,
             "warnings": warnings,
         }
     finally:
@@ -529,6 +720,88 @@ def run(config: BenchmarkConfig) -> dict[str, Any]:
         client.disconnect()
 
 
+def summarize_scaling_trials(
+    trials: Sequence[dict[str, Any]],
+    plateau_threshold_percent: float = 5.0,
+) -> list[dict[str, Any]]:
+    grouped: dict[int, list[float]] = {}
+    for trial in trials:
+        grouped.setdefault(int(trial["parallelism"]), []).append(float(trial["throughput"]))
+    if not grouped:
+        return []
+
+    parallelisms = sorted(grouped)
+    medians = {value: statistics.median(grouped[value]) for value in parallelisms}
+    baseline_parallelism = 1 if 1 in medians else parallelisms[0]
+    baseline = medians[baseline_parallelism]
+    summary = []
+    previous = None
+    for parallelism in parallelisms:
+        values = grouped[parallelism]
+        mean = statistics.fmean(values)
+        stdev = statistics.stdev(values) if len(values) > 1 else 0.0
+        median = medians[parallelism]
+        gain = None if previous is None else (median / previous - 1.0) * 100.0
+        summary.append({
+            "parallelism": parallelism,
+            "trials": len(values),
+            "median_gib_per_second": median,
+            "mean_gib_per_second": mean,
+            "stdev_gib_per_second": stdev,
+            "coefficient_of_variation_percent": stdev / mean * 100.0 if mean else None,
+            "min_gib_per_second": min(values),
+            "max_gib_per_second": max(values),
+            "baseline_parallelism": baseline_parallelism,
+            "speedup_vs_baseline": median / baseline if baseline else None,
+            "speedup_vs_one": median / baseline if baseline and baseline_parallelism == 1 else None,
+            "gain_vs_previous_percent": gain,
+            "plateau": None if gain is None else gain < plateau_threshold_percent,
+        })
+        previous = median
+    return summary
+
+
+def run_sweep(
+    config: BenchmarkConfig,
+    parallelisms: Sequence[int],
+    trials: int,
+    *,
+    plateau_threshold_percent: float = 5.0,
+    runner: Callable[[BenchmarkConfig], dict[str, Any]] = run,
+) -> dict[str, Any]:
+    if trials <= 0:
+        raise ValueError("trials must be positive")
+    values = tuple(dict.fromkeys(int(value) for value in parallelisms))
+    if not values or any(value <= 0 for value in values):
+        raise ValueError("parallelism values must be positive")
+
+    schedule = [(parallelism, trial) for trial in range(trials) for parallelism in values]
+    random.Random(config.seed).shuffle(schedule)
+    raw = []
+    for run_index, (parallelism, trial_index) in enumerate(schedule):
+        table = f"{config.table}_p{parallelism}_t{trial_index}_{uuid.uuid4().hex[:6]}"
+        trial_config = replace(config, parallelism=parallelism, table=table)
+        report = runner(trial_config)
+        throughput = float(report["measurement"]["total"]["logical_gib_per_second"])
+        raw.append({
+            "run_index": run_index,
+            "trial_index": trial_index,
+            "parallelism": parallelism,
+            "throughput": throughput,
+            "report": report,
+        })
+
+    return {
+        "benchmark": "dmi_clickhouse_host_scaling",
+        "base_config": _safe_config(config),
+        "parallelisms": list(values),
+        "trials_per_parallelism": trials,
+        "plateau_threshold_percent": plateau_threshold_percent,
+        "trials": raw,
+        "summary": summarize_scaling_trials(raw, plateau_threshold_percent),
+    }
+
+
 def build_parser() -> argparse.ArgumentParser:
     parser = argparse.ArgumentParser(description=__doc__)
     parser.add_argument("--rows", type=int, default=10_000)
@@ -540,6 +813,9 @@ def build_parser() -> argparse.ArgumentParser:
     parser.add_argument("--latency-samples", type=int, default=10_000)
     parser.add_argument("--warmup-rows", type=int, default=512)
     parser.add_argument("--parallelism", type=int, default=4)
+    parser.add_argument("--parallelism-sweep")
+    parser.add_argument("--trials", type=int, default=1)
+    parser.add_argument("--plateau-threshold-percent", type=float, default=5.0)
     parser.add_argument("--min-batch-bytes", type=parse_byte_size, default=16 * 1024**2)
     parser.add_argument("--max-batch-bytes", type=parse_byte_size, default=64 * 1024**2)
     parser.add_argument("--max-batch-items", type=int, default=10_000)
@@ -550,6 +826,8 @@ def build_parser() -> argparse.ArgumentParser:
     parser.add_argument("--compression", choices=("none", "lz4", "zstd"), default="lz4")
     parser.add_argument("--async-insert", action="store_true")
     parser.add_argument("--drain-timeout-seconds", type=float, default=300.0)
+    parser.add_argument("--socket-timeout-seconds", type=float, default=30.0)
+    parser.add_argument("--server-sample-interval-ms", type=int, default=50)
     parser.add_argument("--host", default="localhost")
     parser.add_argument("--port", type=int, default=9000)
     parser.add_argument("--user", default="default")
@@ -587,6 +865,8 @@ def _config_from_args(args: argparse.Namespace) -> BenchmarkConfig:
         compression=args.compression,
         async_insert=args.async_insert,
         drain_timeout_seconds=args.drain_timeout_seconds,
+        socket_timeout_seconds=args.socket_timeout_seconds,
+        server_sample_interval_ms=args.server_sample_interval_ms,
         host=args.host,
         port=args.port,
         user=args.user,
@@ -602,14 +882,32 @@ def _config_from_args(args: argparse.Namespace) -> BenchmarkConfig:
 def main(argv: Sequence[str] | None = None) -> int:
     args = build_parser().parse_args(argv)
     config = _config_from_args(args)
+    parallelisms = (
+        parse_parallelism_sweep(args.parallelism_sweep)
+        if args.parallelism_sweep
+        else (config.parallelism,)
+    )
+    if len(parallelisms) == 1:
+        config = replace(config, parallelism=parallelisms[0])
+    if args.trials <= 0:
+        raise ValueError("trials must be positive")
     if args.dry_run:
         report: dict[str, Any] = {
             "benchmark": "dmi_clickhouse_host",
             "dry_run": True,
             "config": _safe_config(config),
+            "parallelisms": list(parallelisms),
+            "trials_per_parallelism": args.trials,
             "logical_payload_bytes": config.rows * config.payload_bytes,
             "payload_pool_bytes": config.pool_size * config.payload_bytes,
         }
+    elif len(parallelisms) > 1 or args.trials > 1:
+        report = run_sweep(
+            config,
+            parallelisms,
+            args.trials,
+            plateau_threshold_percent=args.plateau_threshold_percent,
+        )
     else:
         report = run(config)
     output = json.dumps(report, indent=2, sort_keys=True, default=str)
diff --git a/docs/benchmarks.md b/docs/benchmarks.md
index 951c19c74..801b3fd4d 100644
--- a/docs/benchmarks.md
+++ b/docs/benchmarks.md
@@ -30,11 +30,14 @@ make -C native host -j
 
 ```bash
 python -m benchmarks.bench_clickhouse_host \
-  --rows 10000 \
+  --rows 100000 \
   --payload-bytes 64KiB \
+  --parallelism-sweep 1,2,4,8 \
+  --trials 5 \
   --min-batch-bytes 16MiB \
   --max-batch-bytes 64MiB \
   --compression lz4 \
+  --socket-timeout-seconds 30 \
   --json-output host-clickhouse.json
 ```
 
@@ -43,22 +46,53 @@ connecting to ClickHouse. The benchmark creates a uniquely named table and
 drops it after collection; `--keep-table` preserves it for inspection. Set the
 password with `DMI_CLICKHOUSE_PASSWORD` to keep it out of shell history.
 
-The JSON separates enqueue time from total drain time. It also reports logical
-payload throughput, sampled submit latency, DMI process CPU and peak RSS, active
-MergeTree parts, compression size, primary-key size, and insert batching from
-`system.query_log`. Query-log metrics degrade to a warning when the account
-lacks access. The process measurements include the Python-to-C++ call used by
-this synthetic driver; they do not include a ClickHouse server running in a
-different process.
-
-Compare one setting at a time with the same row count, payload pattern, seed,
-and pool size. Useful sweeps are `--parallelism`, batch byte limits,
-`--compression`, and `--async-insert`. Synchronous batching remains the default.
+The scaling report retains every raw trial and summarizes median throughput,
+variance, speedup over the reported baseline, and gain over the preceding
+worker count. `speedup_vs_one` is populated only when the sweep includes one
+worker. Trial order is deterministically shuffled. Startup ends only after
+every native worker has connected and initialized, so steady-state throughput
+does not mix in connection setup.
+
+Each trial reports enqueue and total drain time, sampled submit latency, DMI
+process CPU, process-lifetime peak RSS, per-worker batches/rows/bytes/insert
+time, and peak simultaneous native inserts. The RSS value is not an isolated
+per-trial peak, so do not use later values in an in-process sweep for memory
+scaling. A 50 ms sampler records server-side active inserts,
+query/merge/connection gauges, normalized CPU and I/O wait, resident memory,
+part pressure, and block/network counter deltas over the same steady-state
+interval as throughput. Set
+`--server-sample-interval-ms 0` when the benchmark account cannot read those
+tables. Query-log and server metrics degrade to warnings when unavailable.
+
+Treat saturation as evidence from several signals, not the first flat number:
+throughput gain should fall below the configured plateau threshold across
+repeated trials, realized insert concurrency should reach the requested worker
+count, and client CPU/backpressure plus ClickHouse query/merge metrics should
+identify which side is limiting progress. The process measurements include the
+Python-to-C++ synthetic producer but not ClickHouse server CPU when the server
+runs in another process.
+
+Queue admission uses a finite timeout. The drain deadline is not restarted
+during abort cleanup, and native connect/send/receive calls use the configured
+socket timeout. An in-flight native call cannot be cancelled, so failure
+cleanup can extend past the drain deadline by up to the socket timeout.
+
+Use `--parallelism-sweep` for worker scaling, then compare one additional
+setting at a time with the same row count, payload pattern, seed, and pool size.
+Useful follow-up sweeps are batch byte limits, `--compression`, and
+`--async-insert`. The benchmark explicitly sets `async_insert=0` for
+synchronous trials instead of inheriting the server or user profile; the JSON
+records the effective client settings. This matters on releases such as
+[ClickHouse 26.3](https://clickhouse.com/blog/clickhouse-release-26-03), which
+changed the server default.
 ClickHouse recommends batching synchronous inserts, commonly at least 1,000
 rows and ideally 10,000–100,000 where row size permits; tensor workloads may
 reach practical byte limits earlier. See ClickHouse's
 [insert strategy](https://clickhouse.com/docs/concepts/best-practices/selecting-an-insert-strategy),
 [`system.query_log`](https://clickhouse.com/docs/reference/system-tables/query_log),
+[`system.processes`](https://clickhouse.com/docs/reference/system-tables/processes),
+[`system.metrics`](https://clickhouse.com/docs/reference/system-tables/metrics),
+[`system.asynchronous_metrics`](https://clickhouse.com/docs/reference/system-tables/asynchronous_metrics),
 and [`system.parts`](https://clickhouse.com/docs/reference/system-tables/parts)
 documentation when interpreting results.
 
diff --git a/docs/clickhouse-offload-pipeline.html b/docs/clickhouse-offload-pipeline.html
new file mode 100644
index 000000000..42c084cc9
--- /dev/null
+++ b/docs/clickhouse-offload-pipeline.html
@@ -0,0 +1,153 @@
+<!doctype html>
+<html lang="en">
+<head>
+  <meta charset="utf-8">
+  <meta name="viewport" content="width=device-width, initial-scale=1">
+  <title>DMI ClickHouse offload pipeline</title>
+  <style>
+    :root {
+      color-scheme: light dark;
+      --bg: #09121d;
+      --panel: #122235;
+      --text: #eef7ff;
+      --muted: #a7bdd2;
+      --line: #35526c;
+      --gpu: #8aa7ff;
+      --host: #55d6be;
+      --db: #ffc857;
+      --error: #ff7b8b;
+    }
+    * { box-sizing: border-box; }
+    body {
+      margin: 0;
+      background: radial-gradient(circle at top, #173653, var(--bg) 55%);
+      color: var(--text);
+      font: 16px/1.5 system-ui, sans-serif;
+    }
+    main { width: min(1180px, 94vw); margin: auto; padding: 48px 0 64px; }
+    h1 { margin: 0 0 8px; font-size: clamp(2rem, 5vw, 3.4rem); }
+    h2 { margin: 0 0 8px; font-size: 1.05rem; }
+    p { color: var(--muted); }
+    code { color: var(--db); }
+    .flow {
+      display: grid;
+      grid-template-columns: repeat(5, 1fr);
+      gap: 14px;
+      margin: 36px 0 24px;
+    }
+    .card {
+      min-height: 190px;
+      padding: 20px;
+      border: 1px solid var(--line);
+      border-radius: 16px;
+      background: color-mix(in srgb, var(--panel) 94%, transparent);
+      box-shadow: 0 16px 40px #0004;
+    }
+    .gpu { border-color: var(--gpu); }
+    .host { border-color: var(--host); }
+    .db { border-color: var(--db); }
+    .tag {
+      display: inline-block;
+      margin: 4px 4px 0 0;
+      padding: 3px 8px;
+      border: 1px solid var(--line);
+      border-radius: 999px;
+      color: var(--muted);
+      font-size: .8rem;
+    }
+    .rules { display: grid; grid-template-columns: 1fr 1fr; gap: 14px; }
+    .measure { display: grid; grid-template-columns: repeat(3, 1fr); gap: 14px; margin-top: 14px; }
+    .error { border-color: var(--error); }
+    ul { margin: 10px 0 0; padding-left: 20px; color: var(--muted); }
+    @media (max-width: 900px) {
+      .flow { grid-template-columns: 1fr; }
+      .card { min-height: 0; }
+      .rules { grid-template-columns: 1fr; }
+      .measure { grid-template-columns: 1fr; }
+    }
+  </style>
+</head>
+<body>
+  <main>
+    <h1>GPU capture to ClickHouse</h1>
+    <p>The offload path overlaps GPU capture, host transfer, reconstruction,
+      and native ClickHouse insertion while keeping ownership and failures explicit.</p>
+
+    <section class="flow" aria-label="DMI offload stages">
+      <article class="card gpu">
+        <h2>1. GPU producer</h2>
+        <p>Copies an activation into the payload ring and publishes its descriptor
+          with a system-scope release fence.</p>
+        <span class="tag">device memory</span><span class="tag">CUDA stream</span>
+      </article>
+      <article class="card gpu">
+        <h2>2. Capacity owner</h2>
+        <p>Atomically checks payload, pinned-staging, and task-slot capacity before
+          reserving the next step or eager hook.</p>
+        <span class="tag">one snapshot</span><span class="tag">backpressure</span>
+      </article>
+      <article class="card host">
+        <h2>3. Drain thread</h2>
+        <p>Batches checked D2H copies into pinned memory. Ring space is committed
+          only after the CUDA copy stream succeeds.</p>
+        <span class="tag">pinned ring</span><span class="tag">100 ms flush</span>
+      </article>
+      <article class="card host">
+        <h2>4. P2P thread</h2>
+        <p>Reconstructs pageable tensors, validates metadata, slices requests, and
+          submits rows to the host pipeline.</p>
+        <span class="tag">ATen CPU</span><span class="tag">request slices</span>
+      </article>
+      <article class="card db">
+        <h2>5. ClickHouse stage</h2>
+        <p>Targets 16 MiB batches with a 50 ms linger and bounded queueing. Each
+          stage initializes its own destination once.</p>
+        <span class="tag">native protocol</span><span class="tag">512 MiB queue</span>
+      </article>
+    </section>
+
+    <section class="rules">
+      <article class="card">
+        <h2>Ownership rules</h2>
+        <ul>
+          <li>GPU payload is reusable only after successful D2H completion.</li>
+          <li>Pinned bytes are reusable only after pageable reconstruction.</li>
+          <li>ClickHouse column views retain tensor storage through insertion.</li>
+        </ul>
+      </article>
+      <article class="card error">
+        <h2>Failure rules</h2>
+        <ul>
+          <li>The first CUDA, reconstruction, submission, or worker error wins.</li>
+          <li>New reservations stop after failure.</li>
+          <li>Shutdown cleans every stage, then raises the preserved error.</li>
+        </ul>
+      </article>
+    </section>
+
+    <h2 style="margin-top: 30px">CPU ingestion scaling</h2>
+    <p>Parallelism is useful only when requested workers become ready, perform
+      overlapping inserts, and move the ClickHouse throughput ceiling.</p>
+    <section class="measure" aria-label="CPU ingestion measurement stages">
+      <article class="card host">
+        <h2>Ready boundary</h2>
+        <p>The steady-state clock starts after every worker has connected,
+          selected the database, and applied its session settings.</p>
+        <span class="tag">startup separated</span><span class="tag">finite socket timeout</span>
+      </article>
+      <article class="card host">
+        <h2>Realized concurrency</h2>
+        <p>Snapshots retain per-worker batches, rows, bytes, insert time, and
+          peak simultaneous native inserts.</p>
+        <span class="tag">worker distribution</span><span class="tag">client peak</span>
+      </article>
+      <article class="card db">
+        <h2>Saturation evidence</h2>
+        <p>Repeated shuffled worker sweeps combine throughput variance with
+          sampled active inserts, queries, merges, and connections.</p>
+        <span class="tag">1 → 2 → 4 → 8</span><span class="tag">raw trials retained</span>
+      </article>
+    </section>
+  </main>
+</body>
+</html>
diff --git a/docs/config.md b/docs/config.md
index 7cc673a1c..4d13678a1 100644
--- a/docs/config.md
+++ b/docs/config.md
@@ -38,20 +38,22 @@ ring. Force flush at 100% capacity is always active (prevents deadlock).
 | `drain_flush_payload_ratio` | `float` | 0.5 | Flush when scanned payload bytes >= this fraction of `payload_ring_bytes`. 0 = disabled. |
 | `drain_flush_entry_threshold` | `uint64_t` | 0 | Flush after N entries ready. 0 = disabled. |
 | `drain_flush_byte_threshold` | `uint64_t` | 0 | Flush after N payload bytes ready. 0 = disabled. |
-| `drain_flush_timeout_us` | `uint64_t` | 0 | If a complete tensor has been pending for longer than this many microseconds, flush unconditionally. 0 = disabled. |
+| `drain_flush_timeout_us` | `uint64_t` | 100000 | Flush completed tensors after this many microseconds. 0 disables the timer. |
 
-By default, timeout-based flushing is disabled and the drain thread flushes at
-50% payload-ring usage. If `drain_flush_payload_ratio` and all other thresholds
-are explicitly set to 0, the drain thread only flushes when the ring is 100%
-full or at `stop()` time.
+By default, the drain thread flushes after 100 ms or at 50% payload-ring usage.
+If all thresholds and the timeout are explicitly set to 0, it flushes only when
+the ring is full or at `stop()` time.
 
 ## P2P Thread / Output
 
 | Parameter | Type | Default | Description |
 |-----------|------|---------|-------------|
 | `clone_slices` | `bool` | false | Clone per-request slices before submitting to the host engine. When true (and batch > 1), each slice is an independent tensor so the full assembled tensor can be freed immediately. When false, slices are views that keep the full tensor alive until consumed. |
-| `insert_queue_max_bytes` | `uint64_t` | 4 GiB | ClickHouse insert queue byte limit. The p2p thread blocks when the queue is full. |
-| `insert_queue_max_items` | `uint64_t` | 65536 | ClickHouse insert queue item-count limit. |
+| `insert_queue_max_bytes` | `uint64_t` | 4 GiB | Reserved; does not configure the host queue. |
+| `insert_queue_max_items` | `uint64_t` | 65536 | Reserved; does not configure the host queue. |
+
+Configure ClickHouse batching and backpressure through
+`StageConfig.input_queue`.
 
 ## Constants (not configurable)
 
@@ -59,7 +61,7 @@ full or at `stop()` time.
 |----------|-------|----------|-------------|
 | `PAYLOAD_ALIGN` | 16 bytes | `ring_config.h` | Payload allocation alignment. Every reservation is rounded up to this for vectorized uint4 D2D copies. `payload_ring_bytes` must be a multiple of this. |
 | `READY_SEQ_SENTINEL` | `UINT64_MAX` | `task_entry.h` | Sentinel value for `TaskEntry::ready_seq` (slot not yet published). |
-| `TaskEntry` size | 128 bytes | `task_entry.h` | Fixed slot size, `alignas(128)` for cache-line isolation. |
+| `TaskEntry` size | 64 bytes | `task_entry.h` | Fixed slot size, `alignas(64)` for cache-line isolation. |
 
 ## Python Usage
 
@@ -74,7 +76,6 @@ cfg.drain_poll_timeout_us = 100
 cfg.drain_flush_entry_threshold = 64
 cfg.drain_flush_timeout_us = 1000
 cfg.clone_slices = False
-cfg.insert_queue_max_items = 4096
 
 engine = RingEngine(cfg, host_engine)
 engine.init(stream_handle)
diff --git a/docs/integration-api-v1.md b/docs/integration-api-v1.md
index 36cc7c985..77310d9b8 100644
--- a/docs/integration-api-v1.md
+++ b/docs/integration-api-v1.md
@@ -814,7 +814,7 @@ do not reconfigure it.
 | `drain_flush_payload_ratio` | `0.5` | Payload-capacity flush fraction. |
 | `drain_flush_entry_threshold` | `0` | Absolute ready-entry trigger; zero disables it. |
 | `drain_flush_byte_threshold` | `0` | Absolute ready-byte trigger; zero disables it. |
-| `drain_flush_timeout_us` | `0` | Pending-data age trigger; zero disables it. |
+| `drain_flush_timeout_us` | `100000` | Pending-data age trigger; zero disables it. |
 | `clone_slices` | `False` | Clone multi-request slices so full assembled tensors can be released sooner. |
 | `insert_queue_max_bytes` | `4 GiB` | Reserved field; current v1 does not apply it to host queue limits. |
 | `insert_queue_max_items` | `65536` | Reserved field; current v1 does not apply it to host queue limits. |
@@ -844,12 +844,15 @@ Construct `ClickHouseClientConfig()` and set all fields before passing it to
 | `drop_existing_database` | `False` | Drop the entire configured database before setup. Destructive; isolated tests only. |
 | `client_side_compress` | `none` | `none`, `lz4`, `zstd`, `true`, or `false`. |
 | `index_granularity` | `8192` | MergeTree index granularity for a created table. |
+| `connect_timeout_ms` | `5000` | Native socket connect timeout. |
+| `receive_timeout_ms` | `0` | Native socket receive timeout; `0` is unbounded. |
+| `send_timeout_ms` | `0` | Native socket send timeout; `0` is unbounded. |
 
 Database connection and schema initialization occur asynchronously in stage
 worker threads after host start, not in the config/factory constructor.
-Existing incompatible tables are not migrated. Current v1 also protects schema
-DDL with one process-global one-time guard: use one writer destination per
-process or pre-create any later destination yourself.
+Existing incompatible tables are not migrated. Schema DDL runs once per
+ClickHouse stage, so separate stages may target separate destinations in one
+process.
 
 ### `StageConfig`
 
@@ -874,6 +877,11 @@ The object also exposes mutable `name`, `parallelism`, and
 worker labels. Configure all fields before constructing `DMXHostEngine`; the
 engine copies the stage.
 
+The ClickHouse factory targets 16 MiB batches with a 50 ms linger, caps batches
+at 10,000 rows, and applies backpressure at 20,000 rows or 512 MiB. It leaves
+`max_batch_size` unset so a larger activation can be inserted as a singleton.
+Callers may override these values before constructing `DMXHostEngine`.
+
 ### `QueueConfig`
 
 ```python
@@ -925,6 +933,8 @@ database. Public lifecycle and diagnostics are:
 
 ```python
 start() -> None
+wait_until_ready(timeout_s: float) -> bool
+clickhouse_metrics() -> ClickHouseMetricsSnapshot
 close_input() -> None
 stop(graceful: bool = True, timeout_s: float | None = None) -> bool
 request_abort() -> None
@@ -934,9 +944,24 @@ raise_if_failed() -> None
 ```
 
 `start()` is asynchronous: it can return before a worker fails to connect or
-initialize. `stop()` returning true means threads joined, not that inserts
-succeeded. After shutdown, call `raise_if_failed()`; `failures()` returns
-records with `stage`, `thread_name`, `where`, `exc_type`, and `exc_what`.
+initialize. `wait_until_ready()` returns when every configured ClickHouse
+worker has initialized, a worker fails, or the timeout expires. On false, call
+`raise_if_failed()` to distinguish failure from timeout. `stop()` returning true
+means threads joined, not that inserts succeeded. After shutdown, call
+`raise_if_failed()`; `failures()` returns records with `stage`, `thread_name`,
+`where`, `exc_type`, and `exc_what`.
+
+`clickhouse_metrics()` returns a read-only snapshot with `expected_workers`,
+`ready_workers`, `active_inserts`, `peak_active_inserts`, `batches`, `rows`,
+`logical_bytes`, `insert_seconds`, and a `workers` list. Each worker record has
+`worker_index`, `batches`, `rows`, `logical_bytes`, and `insert_seconds`.
+Insert time covers the synchronous native ClickHouse call; it excludes queue
+wait, column construction, and worker initialization.
+
+Host-engine construction creates fresh schema-initialization and metrics state
+using the stage's final `parallelism`. Mutating a factory-created stage before
+construction therefore keeps readiness and per-worker metrics aligned; engines
+constructed from the same stage do not share accumulated metrics.
 
 ### `ThreadFailure`
 
diff --git a/native/csrc/bindings.cpp b/native/csrc/bindings.cpp
index e4ac9970e..8190cafae 100644
--- a/native/csrc/bindings.cpp
+++ b/native/csrc/bindings.cpp
@@ -54,6 +54,12 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
                      &dmx_host::ClickHouseClientConfig::client_side_compress)
       .def_readwrite("index_granularity",
                      &dmx_host::ClickHouseClientConfig::index_granularity)
+      .def_readwrite("connect_timeout_ms",
+                     &dmx_host::ClickHouseClientConfig::connect_timeout_ms)
+      .def_readwrite("receive_timeout_ms",
+                     &dmx_host::ClickHouseClientConfig::receive_timeout_ms)
+      .def_readwrite("send_timeout_ms",
+                     &dmx_host::ClickHouseClientConfig::send_timeout_ms)
 
       // Expose client_settings as a dict, store internally as unordered_map<string, variant<...>>.
       // This avoids requiring <pybind11/stl_variant.h>.
@@ -107,6 +113,24 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
   using EnqueuePolicy = DMXHostEngine::EnqueuePolicy;
   using Duration = DMXHostEngine::Duration;
 
+  py::class_<dmx_host::ClickHouseWorkerMetrics>(m, "ClickHouseWorkerMetrics")
+      .def_readonly("worker_index", &dmx_host::ClickHouseWorkerMetrics::worker_index)
+      .def_readonly("batches", &dmx_host::ClickHouseWorkerMetrics::batches)
+      .def_readonly("rows", &dmx_host::ClickHouseWorkerMetrics::rows)
+      .def_readonly("logical_bytes", &dmx_host::ClickHouseWorkerMetrics::logical_bytes)
+      .def_readonly("insert_seconds", &dmx_host::ClickHouseWorkerMetrics::insert_seconds);
+
+  py::class_<dmx_host::ClickHouseMetricsSnapshot>(m, "ClickHouseMetricsSnapshot")
+      .def_readonly("expected_workers", &dmx_host::ClickHouseMetricsSnapshot::expected_workers)
+      .def_readonly("ready_workers", &dmx_host::ClickHouseMetricsSnapshot::ready_workers)
+      .def_readonly("active_inserts", &dmx_host::ClickHouseMetricsSnapshot::active_inserts)
+      .def_readonly("peak_active_inserts", &dmx_host::ClickHouseMetricsSnapshot::peak_active_inserts)
+      .def_readonly("batches", &dmx_host::ClickHouseMetricsSnapshot::batches)
+      .def_readonly("rows", &dmx_host::ClickHouseMetricsSnapshot::rows)
+      .def_readonly("logical_bytes", &dmx_host::ClickHouseMetricsSnapshot::logical_bytes)
+      .def_readonly("insert_seconds", &dmx_host::ClickHouseMetricsSnapshot::insert_seconds)
+      .def_readonly("workers", &dmx_host::ClickHouseMetricsSnapshot::workers);
+
   py::enum_<dmx_host::OnFullPolicy>(m, "OnFullPolicy")
       .value("RAISE", dmx_host::OnFullPolicy::RAISE)
       .value("DROP", dmx_host::OnFullPolicy::DROP)
@@ -184,11 +208,17 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
             StageConfig cfg;
             cfg.name = std::move(name);
             cfg.parallelism = parallelism;
+            cfg.input_queue.min_batch_items.reset();
+            cfg.input_queue.min_batch_size = 16ULL * 1024 * 1024;
+            cfg.input_queue.max_linger = Duration(0.05);
+            cfg.input_queue.max_batch_items = 10000;
+            cfg.input_queue.high_watermark_items = 20000;
+            cfg.input_queue.high_watermark_size = 512ULL * 1024 * 1024;
             cfg.process_fn = [](std::vector<dmx_host::dmx_host_queue_item> batch, QueueT* next_q) {
               return dmx_host::ClickHouseInsertStage::ProcessFn<QueueT>(std::move(batch), next_q);
             };
-            // Stored by value in std::any; ClickHouseInsertStage::ThreadInitAny will any_cast it.
-            cfg.thread_init_config = ch_cfg;
+            auto thread_cfg = ch_cfg;
+            cfg.thread_init_config = std::move(thread_cfg);
             cfg.thread_init = &dmx_host::ClickHouseInsertStage::ThreadInitAny;
             cfg.thread_cleanup = &dmx_host::ClickHouseInsertStage::ThreadCleanupAny;
             return cfg;
@@ -200,6 +230,13 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
   py::class_<DMXHostEngine, std::shared_ptr<DMXHostEngine>>(m, "DMXHostEngine")
       .def(py::init<StageConfig>(), py::arg("insert_stage"))
       .def("start", &DMXHostEngine::start)
+      .def("wait_until_ready",
+           [](DMXHostEngine& self, double timeout_s) {
+             return self.wait_until_ready(DMXHostEngine::Duration(timeout_s));
+           },
+           py::arg("timeout_s"),
+           py::call_guard<py::gil_scoped_release>())
+      .def("clickhouse_metrics", &DMXHostEngine::clickhouse_metrics)
       .def("stop",
            [](DMXHostEngine& self, bool graceful, std::optional<double> timeout_s) {
              if (timeout_s) {
@@ -314,13 +351,12 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
       .def("staging_cap", &ring_py::RingEnginePy::staging_cap)
       .def("task_cap",    &ring_py::RingEnginePy::task_cap)
       .def("payload_tensor", &ring_py::RingEnginePy::payload_tensor)
-      // Safety-net surface (eager only).  available_capacity() and
-      // reserve_one() are CPU-only and fast -- no GIL release needed.
+      // Safety-net surface (eager only). CPU-only and fast.
       // flush_and_wait() blocks on cudaStreamSynchronize + drain flush --
       // GIL released so other Python threads aren't blocked.
-      .def("available_capacity", &ring_py::RingEnginePy::available_capacity)
-      .def("reserve_one",
-           &ring_py::RingEnginePy::reserve_one,
+      .def("effective_capacity", &ring_py::RingEnginePy::effective_capacity)
+      .def("try_reserve_one",
+           &ring_py::RingEnginePy::try_reserve_one,
            py::arg("nbytes"))
       .def("flush_and_wait",
            &ring_py::RingEnginePy::flush_and_wait,
diff --git a/native/csrc/clickhouse_client.cpp b/native/csrc/clickhouse_client.cpp
index 454aa2901..56f4815c1 100644
--- a/native/csrc/clickhouse_client.cpp
+++ b/native/csrc/clickhouse_client.cpp
@@ -7,6 +7,7 @@
 #include <algorithm>
 #include <cctype>
 #include <charconv>
+#include <chrono>
 #include <mutex>
 #include <sstream>
 #include <stdexcept>
@@ -19,6 +20,75 @@
 #include <clickhouse/types/types.h>
 
 namespace dmx_host {
+
+ClickHouseRuntimeMetrics::ClickHouseRuntimeMetrics(int expected_workers)
+    : expected_workers_(expected_workers),
+      ready_(static_cast<std::size_t>(std::max(0, expected_workers)), false),
+      workers_(static_cast<std::size_t>(std::max(0, expected_workers))) {
+  if (expected_workers < 0) {
+    throw std::invalid_argument("expected_workers must be non-negative");
+  }
+  for (int i = 0; i < expected_workers; ++i) workers_[i].worker_index = i;
+}
+
+void ClickHouseRuntimeMetrics::WorkerReady(int worker_index) {
+  std::lock_guard<std::mutex> lock(mu_);
+  if (worker_index < 0 || worker_index >= expected_workers_) {
+    throw std::out_of_range("ClickHouse worker index is out of range");
+  }
+  if (!ready_[worker_index]) {
+    ready_[worker_index] = true;
+    ++ready_workers_;
+    ready_cv_.notify_all();
+  }
+}
+
+bool ClickHouseRuntimeMetrics::WaitUntilReady(std::chrono::milliseconds timeout) {
+  std::unique_lock<std::mutex> lock(mu_);
+  return ready_cv_.wait_for(lock, timeout, [&] {
+    return ready_workers_ == expected_workers_;
+  });
+}
+
+void ClickHouseRuntimeMetrics::BeginInsert() {
+  std::lock_guard<std::mutex> lock(mu_);
+  ++active_inserts_;
+  peak_active_inserts_ = std::max(peak_active_inserts_, active_inserts_);
+}
+
+void ClickHouseRuntimeMetrics::EndInsert(int worker_index, std::uint64_t rows,
+                                         std::uint64_t logical_bytes,
+                                         double seconds) {
+  std::lock_guard<std::mutex> lock(mu_);
+  if (active_inserts_ > 0) --active_inserts_;
+  ++batches_;
+  rows_ += rows;
+  logical_bytes_ += logical_bytes;
+  insert_seconds_ += seconds;
+  if (worker_index >= 0 && worker_index < expected_workers_) {
+    auto& worker = workers_[worker_index];
+    ++worker.batches;
+    worker.rows += rows;
+    worker.logical_bytes += logical_bytes;
+    worker.insert_seconds += seconds;
+  }
+}
+
+ClickHouseMetricsSnapshot ClickHouseRuntimeMetrics::Snapshot() const {
+  std::lock_guard<std::mutex> lock(mu_);
+  ClickHouseMetricsSnapshot snapshot;
+  snapshot.expected_workers = expected_workers_;
+  snapshot.ready_workers = ready_workers_;
+  snapshot.active_inserts = active_inserts_;
+  snapshot.peak_active_inserts = peak_active_inserts_;
+  snapshot.batches = batches_;
+  snapshot.rows = rows_;
+  snapshot.logical_bytes = logical_bytes_;
+  snapshot.insert_seconds = insert_seconds_;
+  snapshot.workers = workers_;
+  return snapshot;
+}
+
 namespace {
 
 thread_local std::unique_ptr<clickhouse::Client> tl_client;
@@ -26,8 +96,8 @@ thread_local bool tl_inited = false;
 thread_local bool tl_cleaned = false;
 thread_local std::string tl_db;
 thread_local std::string tl_table;
-
-std::once_flag g_schema_once;
+thread_local int tl_worker_index = -1;
+thread_local std::shared_ptr<ClickHouseRuntimeMetrics> tl_runtime_metrics;
 
 // --------------------- SQL helpers ---------------------
 
@@ -377,7 +447,7 @@ StagedRow StageOneRow(ClickHouseRow&& row) {
 
 // ===================== Stage API =====================
 
-void ClickHouseInsertStage::ThreadInit(int /*thread_idx*/, const ClickHouseClientConfig& cfg) {
+void ClickHouseInsertStage::ThreadInit(int thread_idx, const ClickHouseClientConfig& cfg) {
   if (tl_inited) {
     throw std::runtime_error("ClickHouseInsertStage::ThreadInit() called more than once in the same thread");
   }
@@ -393,6 +463,13 @@ void ClickHouseInsertStage::ThreadInit(int /*thread_idx*/, const ClickHouseClien
   opts.SetPort(static_cast<uint16_t>(cfg.port));
   opts.SetUser(cfg.username);
   opts.SetPassword(cfg.password);
+  if (cfg.connect_timeout_ms < 0 || cfg.receive_timeout_ms < 0 ||
+      cfg.send_timeout_ms < 0) {
+    throw std::invalid_argument("ClickHouse socket timeouts must be non-negative");
+  }
+  opts.SetConnectionConnectTimeout(std::chrono::milliseconds(cfg.connect_timeout_ms));
+  opts.SetConnectionRecvTimeout(std::chrono::milliseconds(cfg.receive_timeout_ms));
+  opts.SetConnectionSendTimeout(std::chrono::milliseconds(cfg.send_timeout_ms));
 
   // stable default database for handshake
   opts.SetDefaultDatabase("default");
@@ -416,8 +493,8 @@ void ClickHouseInsertStage::ThreadInit(int /*thread_idx*/, const ClickHouseClien
   }
 
   try {
-    // DDL init once globally (copy cfg/opts into the call_once closure)
-    std::call_once(g_schema_once, [cfg, opts]() { RunSchemaInitOnce(cfg, opts); });
+    std::call_once(*cfg.schema_once,
+                   [cfg, opts]() { RunSchemaInitOnce(cfg, opts); });
 
     // Per-thread client
     tl_client = std::make_unique<clickhouse::Client>(opts);
@@ -435,19 +512,28 @@ void ClickHouseInsertStage::ThreadInit(int /*thread_idx*/, const ClickHouseClien
     throw;
   }
 
+  tl_worker_index = thread_idx;
+  tl_runtime_metrics = cfg.runtime_metrics;
   tl_inited = true;
+  tl_runtime_metrics->WorkerReady(thread_idx);
 }
 
 void ClickHouseInsertStage::ThreadCleanup() noexcept {
   if (!tl_inited || tl_cleaned) return;
   tl_cleaned = true;
   try { tl_client.reset(); } catch (...) {}
+  tl_runtime_metrics.reset();
+  tl_worker_index = -1;
 }
 
 void ClickHouseInsertStage::InsertBatch(std::vector<dmx_host_queue_item>&& batch) {
   clickhouse::Client& client = ClientOrThrow();
   if (batch.empty()) return;
 
+  std::uint64_t logical_bytes = 0;
+  for (const auto& item : batch) logical_bytes += item.item_size;
+  const std::uint64_t row_count = batch.size();
+
   // Stage rows (keeps tensor memory alive for AppendNoManagedLifetime)
   std::vector<StagedRow> rows;
   rows.reserve(batch.size());
@@ -511,7 +597,19 @@ void ClickHouseInsertStage::InsertBatch(std::vector<dmx_host_queue_item>&& batch
   block.AppendColumn("shape", col_shape);
   block.AppendColumn("bytes", col_bytes);
 
-  client.Insert(fq_table, block);
+  const auto started = std::chrono::steady_clock::now();
+  tl_runtime_metrics->BeginInsert();
+  try {
+    client.Insert(fq_table, block);
+  } catch (...) {
+    const double seconds = std::chrono::duration<double>(
+        std::chrono::steady_clock::now() - started).count();
+    tl_runtime_metrics->EndInsert(tl_worker_index, row_count, logical_bytes, seconds);
+    throw;
+  }
+  const double seconds = std::chrono::duration<double>(
+      std::chrono::steady_clock::now() - started).count();
+  tl_runtime_metrics->EndInsert(tl_worker_index, row_count, logical_bytes, seconds);
 }
 
 // Force emission of the specialization referenced from bindings.cpp:
diff --git a/native/csrc/clickhouse_client.h b/native/csrc/clickhouse_client.h
index bf0e2ba0d..6093d4ef7 100644
--- a/native/csrc/clickhouse_client.h
+++ b/native/csrc/clickhouse_client.h
@@ -2,7 +2,11 @@
 #define DMX_HOST_CLICKHOUSE_CLIENT_H_
 
 #include <any>
+#include <chrono>
+#include <condition_variable>
 #include <cstdint>
+#include <memory>
+#include <mutex>
 #include <optional>
 #include <string>
 #include <unordered_map>
@@ -19,6 +23,52 @@ namespace dmx_host {
 // Settings value types for session SET ... (bound from Python too).
 using ClickHouseSettingValue = std::variant<std::string, std::int64_t, bool>;
 
+struct ClickHouseWorkerMetrics {
+  int worker_index = 0;
+  std::uint64_t batches = 0;
+  std::uint64_t rows = 0;
+  std::uint64_t logical_bytes = 0;
+  double insert_seconds = 0.0;
+};
+
+struct ClickHouseMetricsSnapshot {
+  int expected_workers = 0;
+  int ready_workers = 0;
+  std::uint64_t active_inserts = 0;
+  std::uint64_t peak_active_inserts = 0;
+  std::uint64_t batches = 0;
+  std::uint64_t rows = 0;
+  std::uint64_t logical_bytes = 0;
+  double insert_seconds = 0.0;
+  std::vector<ClickHouseWorkerMetrics> workers;
+};
+
+class ClickHouseRuntimeMetrics {
+ public:
+  explicit ClickHouseRuntimeMetrics(int expected_workers = 0);
+
+  void WorkerReady(int worker_index);
+  bool WaitUntilReady(std::chrono::milliseconds timeout);
+  void BeginInsert();
+  void EndInsert(int worker_index, std::uint64_t rows,
+                 std::uint64_t logical_bytes, double seconds);
+  ClickHouseMetricsSnapshot Snapshot() const;
+
+ private:
+  mutable std::mutex mu_;
+  std::condition_variable ready_cv_;
+  int expected_workers_ = 0;
+  int ready_workers_ = 0;
+  std::uint64_t active_inserts_ = 0;
+  std::uint64_t peak_active_inserts_ = 0;
+  std::uint64_t batches_ = 0;
+  std::uint64_t rows_ = 0;
+  std::uint64_t logical_bytes_ = 0;
+  double insert_seconds_ = 0.0;
+  std::vector<bool> ready_;
+  std::vector<ClickHouseWorkerMetrics> workers_;
+};
+
 /**
  * ClickHouse connection + schema init configuration.
  */
@@ -43,6 +93,14 @@ struct ClickHouseClientConfig {
   std::string client_side_compress = "none";
 
   int index_granularity = 8192;
+  int connect_timeout_ms = 5000;
+  int receive_timeout_ms = 0;
+  int send_timeout_ms = 0;
+
+  std::shared_ptr<std::once_flag> schema_once =
+      std::make_shared<std::once_flag>();
+  std::shared_ptr<ClickHouseRuntimeMetrics> runtime_metrics =
+      std::make_shared<ClickHouseRuntimeMetrics>();
 };
 
 /**
diff --git a/native/csrc/dmx_host_engine.h b/native/csrc/dmx_host_engine.h
index f424b9114..11f63da44 100644
--- a/native/csrc/dmx_host_engine.h
+++ b/native/csrc/dmx_host_engine.h
@@ -1,18 +1,49 @@
 #ifndef DMX_HOST_ENGINE__
 #define DMX_HOST_ENGINE__
 
+#include "clickhouse_client.h"
 #include "dmx_host_utils.h"
 #include "pipelined_engine.hpp"
 
+#include <algorithm>
+#include <chrono>
+
 namespace dmx_host{
 
-// DMXHostEngine is a single-stage ClickHouse insert pipeline.
-// Pre-assembled ClickHouseRows are submitted via submit_direct().
-class DMXHostEngine : public PipelinedEngine<dmx_host_queue_item, uint64_t, 1, QueueOptions<false, false, false>, false,
-NoOutputHandler<dmx_host_queue_item> >{
+using DMXHostEngineBase = PipelinedEngine<
+    dmx_host_queue_item, uint64_t, 1, QueueOptions<false, false, false>, false,
+    NoOutputHandler<dmx_host_queue_item>>;
+
+class DMXHostEngine : public DMXHostEngineBase {
 public:
-    explicit DMXHostEngine(StageConfig insert_stage):
-    PipelinedEngine(std::array<StageConfig, 1>{std::move(insert_stage)}, EngineConfig{}){}
+    using Base = DMXHostEngineBase;
+    using StageConfig = Base::StageConfig;
+    using EngineConfig = Base::EngineConfig;
+    using Duration = Base::Duration;
+
+    explicit DMXHostEngine(StageConfig insert_stage)
+        : DMXHostEngine(Prepare(std::move(insert_stage))) {}
+
+    bool wait_until_ready(Duration timeout) {
+        if (timeout.count() < 0.0) {
+            throw std::invalid_argument("timeout must be non-negative");
+        }
+        const auto deadline = std::chrono::steady_clock::now() +
+            std::chrono::duration_cast<std::chrono::steady_clock::duration>(timeout);
+        while (std::chrono::steady_clock::now() < deadline) {
+            const auto remaining = std::chrono::duration_cast<std::chrono::milliseconds>(
+                deadline - std::chrono::steady_clock::now());
+            const auto slice = std::min(remaining, std::chrono::milliseconds(50));
+            if (metrics_->WaitUntilReady(slice)) return true;
+            if (!failures().empty()) return false;
+        }
+        const auto snapshot = metrics_->Snapshot();
+        return snapshot.ready_workers == snapshot.expected_workers;
+    }
+
+    ClickHouseMetricsSnapshot clickhouse_metrics() const {
+        return metrics_->Snapshot();
+    }
 
     // Submit a pre-assembled ClickHouseRow directly to the insert stage.
     // Fields must match the order expected by ClickHouseInsertStage:
@@ -29,6 +60,32 @@ NoOutputHandler<dmx_host_queue_item> >{
         items.emplace_back(std::move(row), nbytes);
         submit_items(std::move(items));
     }
+
+private:
+    struct PreparedStage {
+        StageConfig stage;
+        std::shared_ptr<ClickHouseRuntimeMetrics> metrics;
+    };
+
+    static PreparedStage Prepare(StageConfig stage) {
+        if (stage.parallelism <= 0) {
+            throw std::invalid_argument("parallelism must be positive");
+        }
+        auto config =
+            std::any_cast<ClickHouseClientConfig>(stage.thread_init_config);
+        config.schema_once = std::make_shared<std::once_flag>();
+        auto metrics =
+            std::make_shared<ClickHouseRuntimeMetrics>(stage.parallelism);
+        config.runtime_metrics = metrics;
+        stage.thread_init_config = std::move(config);
+        return PreparedStage{std::move(stage), std::move(metrics)};
+    }
+
+    explicit DMXHostEngine(PreparedStage prepared)
+        : Base(std::array<StageConfig, 1>{std::move(prepared.stage)}, EngineConfig{}),
+          metrics_(std::move(prepared.metrics)) {}
+
+    std::shared_ptr<ClickHouseRuntimeMetrics> metrics_;
 };
 
 }
diff --git a/native/csrc/ring/drain_thread.cpp b/native/csrc/ring/drain_thread.cpp
index 6fbcfa428..a6f397a08 100644
--- a/native/csrc/ring/drain_thread.cpp
+++ b/native/csrc/ring/drain_thread.cpp
@@ -34,16 +34,23 @@ DrainThread::~DrainThread() noexcept {
 
 void DrainThread::start() {
     running_.store(true, std::memory_order_relaxed);
-    thread_ = std::thread([this] { loop(); });
+    thread_ = std::thread([this] {
+        try {
+            loop();
+        } catch (...) {
+            report_failure(std::current_exception());
+        }
+    });
 }
 
 void DrainThread::stop() {
-    if (!running_.exchange(false)) return;
+    running_.store(false, std::memory_order_relaxed);
     cv_.notify_all();
     if (thread_.joinable()) thread_.join();
 }
 
 void DrainThread::notify() {
+    rethrow_if_failed();
     {
         std::lock_guard<std::mutex> lk(mu_);
         notified_ = true;
@@ -58,6 +65,7 @@ void DrainThread::notify() {
 // cudaStreamSynchronize(main_stream) first so all GPU writes are visible.
 // ---------------------------------------------------------------------------
 void DrainThread::force_flush_and_wait() {
+    rethrow_if_failed();
     {
         std::lock_guard<std::mutex> lk(mu_);
         flush_requested_ = true;
@@ -68,7 +76,37 @@ void DrainThread::force_flush_and_wait() {
 
     // Block until drain thread completes the flush
     std::unique_lock<std::mutex> lk(mu_);
-    flush_done_cv_.wait(lk, [this] { return flush_done_; });
+    flush_done_cv_.wait(lk, [this] { return flush_done_ || has_failed(); });
+    lk.unlock();
+    rethrow_if_failed();
+}
+
+void DrainThread::report_failure(std::exception_ptr failure) noexcept {
+    {
+        std::lock_guard<std::mutex> lk(failure_mu_);
+        if (!failure_) failure_ = std::move(failure);
+    }
+    failed_.store(true, std::memory_order_release);
+    running_.store(false, std::memory_order_relaxed);
+    {
+        std::lock_guard<std::mutex> lk(mu_);
+        flush_done_ = true;
+    }
+    {
+        std::lock_guard<std::mutex> lk(pop_mu_);
+        p2p_stop_requested_ = true;
+    }
+    cv_.notify_all();
+    flush_done_cv_.notify_all();
+    pop_cv_.notify_all();
+    staging_cv_.notify_all();
+}
+
+void DrainThread::rethrow_if_failed() const {
+    if (!has_failed()) return;
+    std::lock_guard<std::mutex> lk(failure_mu_);
+    if (failure_) std::rethrow_exception(failure_);
+    throw std::runtime_error("DrainThread failed");
 }
 
 // ---------------------------------------------------------------------------
@@ -111,22 +149,37 @@ void DrainThread::notify_staging_freed_bytes(uint64_t nbytes) {
 }
 
 // ---------------------------------------------------------------------------
-// Capacity query accessors
+// Capacity query
 // ---------------------------------------------------------------------------
-uint64_t DrainThread::cpu_payload_head() const {
-    return cpu_payload_head_;
-}
-
-uint64_t DrainThread::cpu_payload_tail_committed() const {
-    return cpu_payload_tail_committed_;
-}
-
-uint64_t DrainThread::cpu_task_head() const {
-    return cpu_task_head_;
+CapacitySnapshot DrainThread::capacity_snapshot() {
+    rethrow_if_failed();
+    std::lock_guard<std::mutex> lk(mgmt_mu_);
+    return {
+        cpu_payload_head_,
+        cpu_payload_tail_committed_,
+        cpu_task_head_,
+        cpu_task_tail_,
+    };
 }
 
-uint64_t DrainThread::cpu_task_tail_committed() const {
-    return cpu_task_tail_;
+bool DrainThread::try_reserve(uint64_t payload_bytes, uint32_t num_tasks,
+                              uint64_t payload_capacity,
+                              uint64_t task_capacity) {
+    rethrow_if_failed();
+    std::lock_guard<std::mutex> lk(mgmt_mu_);
+    const uint64_t payload_used =
+        cpu_payload_head_ - cpu_payload_tail_committed_;
+    const uint64_t tasks_used = cpu_task_head_ - cpu_task_tail_;
+    if (payload_used > payload_capacity || tasks_used > task_capacity) {
+        throw std::logic_error("DrainThread: capacity invariant violated");
+    }
+    if (payload_bytes > payload_capacity - payload_used ||
+        num_tasks > task_capacity - tasks_used) {
+        return false;
+    }
+    cpu_payload_head_ += payload_bytes;
+    cpu_task_head_ += num_tasks;
+    return true;
 }
 
 // ---------------------------------------------------------------------------
@@ -134,6 +187,7 @@ uint64_t DrainThread::cpu_task_tail_committed() const {
 // Called from prepare_step after confirming space is available.
 // ---------------------------------------------------------------------------
 void DrainThread::reserve(uint64_t payload_bytes, uint32_t num_tasks) {
+    rethrow_if_failed();
     std::lock_guard<std::mutex> lk(mgmt_mu_);
     cpu_payload_head_ += payload_bytes;
     cpu_task_head_    += num_tasks;
@@ -143,6 +197,7 @@ void DrainThread::reserve(uint64_t payload_bytes, uint32_t num_tasks) {
 // submit_cpu_direct -- submit a CPU-direct tensor to drain -> p2p pipeline.
 // ---------------------------------------------------------------------------
 void DrainThread::submit_cpu_direct(at::Tensor cpu_tensor, uint64_t tensor_bytes) {
+    rethrow_if_failed();
     DrainTask task{};
     task.tensor_total_bytes = tensor_bytes;
     task.cpu_paged_tensor   = std::move(cpu_tensor);
@@ -163,7 +218,7 @@ void DrainThread::submit_cpu_direct(at::Tensor cpu_tensor, uint64_t tensor_bytes
 // ---------------------------------------------------------------------------
 void DrainThread::do_full_flush() {
     for (;;) {
-        uint64_t flush_count = 0, flush_bytes = 0;
+        uint64_t flush_count = 0, flush_bytes = 0, src_start = 0;
         {
             std::lock_guard<std::mutex> lk(mgmt_mu_);
             scan_ready();
@@ -175,17 +230,20 @@ void DrainThread::do_full_flush() {
                 flush_count++;
             }
             if (flush_count == 0) break;
-            flush_state_update(flush_count, flush_bytes);
+            src_start = cpu_payload_tail_;
         }
         {
             std::unique_lock<std::mutex> lk(staging_mu_);
-            staging_cv_.wait(lk, [&] { return staging_.free_bytes() >= flush_bytes; });
+            staging_cv_.wait(lk, [&] {
+                return staging_.free_bytes() >= flush_bytes || has_failed();
+            });
         }
-        enqueue_d2h(flush_bytes);
+        rethrow_if_failed();
+        enqueue_d2h(flush_bytes, src_start);
         sync_stream();
         {
             std::lock_guard<std::mutex> lk(mgmt_mu_);
-            cpu_payload_tail_committed_ = cpu_payload_tail_;
+            flush_state_update(flush_count, flush_bytes);
         }
         submit_to_p2p(flush_count, flush_bytes);
         {
@@ -227,7 +285,7 @@ void DrainThread::loop() {
             continue;  // skip normal sleep, re-check immediately
         }
 
-        uint64_t flush_count = 0, flush_bytes = 0;
+        uint64_t flush_count = 0, flush_bytes = 0, src_start = 0;
         bool needs_flush = false;
 
         {
@@ -248,7 +306,7 @@ void DrainThread::loop() {
                             (unsigned long)flush_bytes, (unsigned long)pending_entries_,
                             (unsigned long)staging_.free_bytes());
 
-                    flush_state_update(flush_count, flush_bytes);
+                    src_start = cpu_payload_tail_;
                     needs_flush = true;
                 }
             }
@@ -258,16 +316,17 @@ void DrainThread::loop() {
             {
                 std::unique_lock<std::mutex> lk(staging_mu_);
                 staging_cv_.wait(lk, [&] {
-                    return staging_.free_bytes() >= flush_bytes;
+                    return staging_.free_bytes() >= flush_bytes || has_failed();
                 });
             }
 
-            enqueue_d2h(flush_bytes);
+            rethrow_if_failed();
+            enqueue_d2h(flush_bytes, src_start);
             sync_stream();
 
             {
                 std::lock_guard<std::mutex> lk(mgmt_mu_);
-                cpu_payload_tail_committed_ = cpu_payload_tail_;
+                flush_state_update(flush_count, flush_bytes);
             }
 
             submit_to_p2p(flush_count, flush_bytes);
@@ -289,7 +348,12 @@ void DrainThread::loop() {
     }
 
     // Final flush
-    cudaDeviceSynchronize();
+    cudaError_t error = cudaDeviceSynchronize();
+    if (error != cudaSuccess) {
+        throw std::runtime_error(
+            std::string("DrainThread: cudaDeviceSynchronize failed: ") +
+            cudaGetErrorString(error));
+    }
     do_full_flush();
 }
 
@@ -354,18 +418,23 @@ void DrainThread::flush_state_update(uint64_t flush_count, uint64_t flush_bytes)
         ++cpu_task_tail_;
     }
     cpu_payload_tail_ += flush_bytes;
+    cpu_payload_tail_committed_ = cpu_payload_tail_;
 }
 
 void DrainThread::sync_stream() {
-    cudaStreamSynchronize(stream_);
+    cudaError_t error = cudaStreamSynchronize(stream_);
+    if (error != cudaSuccess) {
+        throw std::runtime_error(
+            std::string("DrainThread: cudaStreamSynchronize failed: ") +
+            cudaGetErrorString(error));
+    }
 }
 
 // ---------------------------------------------------------------------------
-void DrainThread::enqueue_d2h(uint64_t flush_bytes) {
+void DrainThread::enqueue_d2h(uint64_t flush_bytes, uint64_t src_start) {
     if (flush_bytes == 0) return;
     const uint64_t gpu_cap = ring_.payload_cap;
     const uint64_t stg_cap = staging_.capacity();
-    uint64_t src_start = cpu_payload_tail_ - flush_bytes;
     uint64_t gpu_cursor = src_start % gpu_cap;
     uint64_t stg_cursor = staging_.head() % stg_cap;
     uint64_t remaining  = flush_bytes;
@@ -384,8 +453,9 @@ void DrainThread::enqueue_d2h(uint64_t flush_bytes) {
                         ring_.payload_buf + gpu_cursor,
                         chunk, cudaMemcpyDeviceToHost, stream_);
         if (err != cudaSuccess) {
-            RING_DBG("[enqueue_d2h] cudaMemcpyAsync FAILED: %s\n",
-                    cudaGetErrorString(err));
+            throw std::runtime_error(
+                std::string("DrainThread: cudaMemcpyAsync failed: ") +
+                cudaGetErrorString(err));
         }
         RING_DBG("[enqueue_d2h] chunk=%d enqueued OK\n", chunk_idx);
 
@@ -402,7 +472,13 @@ void DrainThread::enqueue_d2h(uint64_t flush_bytes) {
 // ---------------------------------------------------------------------------
 void DrainThread::submit_to_p2p(uint64_t flush_count, uint64_t flush_bytes) {
     uint64_t cumulative = 0;
-    const uint64_t staging_batch_start = staging_.head();
+    uint64_t staging_batch_start = 0;
+    {
+        std::lock_guard<std::mutex> lk(staging_mu_);
+        staging_batch_start = staging_.head();
+    }
+    std::vector<DrainTask> ready;
+    ready.reserve(flush_count);
 
     for (uint64_t i = 0; i < flush_count; ++i) {
         const TaskEntry& ec = scanned_[i];
@@ -430,18 +506,22 @@ void DrainThread::submit_to_p2p(uint64_t flush_count, uint64_t flush_bytes) {
             cumulative += alloc;
         }
 
-        {
-            std::lock_guard<std::mutex> lk(queue_mu_);
-            task_queue_.push_back(std::move(task));
-        }
-        {
-            std::lock_guard<std::mutex> lk(pop_mu_);
-            can_pop_count_ += 1;
-        }
-        pop_cv_.notify_one();
+        ready.push_back(std::move(task));
     }
 
-    staging_.advance_head(flush_bytes);
+    {
+        std::lock_guard<std::mutex> lk(queue_mu_);
+        for (auto& task : ready) task_queue_.push_back(std::move(task));
+    }
+    {
+        std::lock_guard<std::mutex> lk(staging_mu_);
+        staging_.advance_head(flush_bytes);
+    }
+    {
+        std::lock_guard<std::mutex> lk(pop_mu_);
+        can_pop_count_ += flush_count;
+    }
+    pop_cv_.notify_one();
 }
 
 // ---------------------------------------------------------------------------
diff --git a/native/csrc/ring/drain_thread.h b/native/csrc/ring/drain_thread.h
index 66d03780b..d0c1071fa 100644
--- a/native/csrc/ring/drain_thread.h
+++ b/native/csrc/ring/drain_thread.h
@@ -22,12 +22,20 @@
 #include <atomic>
 #include <condition_variable>
 #include <deque>
+#include <exception>
 #include <mutex>
 #include <thread>
 #include <vector>
 
 namespace ring {
 
+struct CapacitySnapshot {
+    uint64_t payload_head;
+    uint64_t payload_tail_committed;
+    uint64_t task_head;
+    uint64_t task_tail_committed;
+};
+
 class DrainThread {
 public:
     DrainThread(RingState& rs, PinnedStaging& staging, const RingConfig& cfg);
@@ -57,13 +65,18 @@ class DrainThread {
 
     void notify_staging_freed_bytes(uint64_t nbytes);
 
+    void report_failure(std::exception_ptr failure) noexcept;
+    void rethrow_if_failed() const;
+    bool has_failed() const noexcept {
+        return failed_.load(std::memory_order_acquire);
+    }
+
     bool is_running() const { return running_.load(std::memory_order_relaxed); }
 
-    // Capacity query accessors (called from RingEnginePy::prepare_step).
-    uint64_t cpu_payload_head() const;
-    uint64_t cpu_payload_tail_committed() const;
-    uint64_t cpu_task_head() const;
-    uint64_t cpu_task_tail_committed() const;
+    CapacitySnapshot capacity_snapshot();
+
+    bool try_reserve(uint64_t payload_bytes, uint32_t num_tasks,
+                     uint64_t payload_capacity, uint64_t task_capacity);
 
     // Pre-allocate ring space for the next step's producer kernels.
     // Advances cpu_payload_head_ and cpu_task_head_ under mgmt_mu_.
@@ -115,6 +128,10 @@ class DrainThread {
     std::mutex              staging_mu_;
     std::condition_variable staging_cv_;
 
+    mutable std::mutex      failure_mu_;
+    std::exception_ptr      failure_;
+    std::atomic<bool>       failed_{false};
+
     void loop();
 
     // Drain all pending entries -- called by the drain thread when
@@ -127,7 +144,7 @@ class DrainThread {
     void flush_state_update(uint64_t flush_count, uint64_t flush_bytes);
 
     void sync_stream();
-    void enqueue_d2h(uint64_t flush_bytes);
+    void enqueue_d2h(uint64_t flush_bytes, uint64_t src_start);
 
     // Split into two: submit_to_p2p pushes DrainTasks to the p2p queue
     // (uses queue_mu_/pop_mu_, NOT mgmt_mu_).  trim_scanned updates
diff --git a/native/csrc/ring/p2p_thread.cpp b/native/csrc/ring/p2p_thread.cpp
index f68426297..60e4beefe 100644
--- a/native/csrc/ring/p2p_thread.cpp
+++ b/native/csrc/ring/p2p_thread.cpp
@@ -6,10 +6,9 @@
 #include "pinned_staging.h"
 
 #include <ATen/ATen.h>
-#include <cstdio>
 #include <cstring>
 #include <exception>
-#include <mutex>
+#include <stdexcept>
 
 namespace ring {
 
@@ -17,36 +16,6 @@ namespace ring {
 // ATen helpers (no GIL required for CPU tensors)
 // ---------------------------------------------------------------------------
 
-static std::once_flag g_submit_failure_log_once;
-
-static void log_submit_failure_once(
-    const std::string& model_id,
-    const std::string& req_id,
-    const std::string& act_name,
-    int32_t layer_no,
-    int32_t shard_rank,
-    int32_t start_token,
-    int32_t end_token,
-    const char* error)
-{
-    std::call_once(g_submit_failure_log_once, [&] {
-        fprintf(stderr,
-                "[DMI][P2P] WARN: failed to submit tensor slice to host "
-                "engine; suppressing further submit errors. model_id=%s "
-                "request_id=%s act_name=%s layer_no=%d shard_rank=%d "
-                "token_range=[%d,%d) error=\"%s\"\n",
-                model_id.c_str(),
-                req_id.c_str(),
-                act_name.c_str(),
-                layer_no,
-                shard_rank,
-                start_token,
-                end_token,
-                error ? error : "unknown");
-        fflush(stderr);
-    });
-}
-
 // Build ClickHouse act_name from hook_type.
 // Per-layer: "blocks.<hook_type_name>"  (e.g. "blocks.attn.hook_pattern")
 // Global:    "<hook_type_name>"         (e.g. "hook_embed", "token_ids")
@@ -132,7 +101,13 @@ P2PThread::~P2PThread() noexcept {
 }
 
 void P2PThread::start() {
-    thread_ = std::thread([this] { loop(); });
+    thread_ = std::thread([this] {
+        try {
+            loop();
+        } catch (...) {
+            drain_.report_failure(std::current_exception());
+        }
+    });
 }
 
 void P2PThread::stop() {
@@ -190,14 +165,16 @@ void P2PThread::process(std::vector<DrainTask>& tasks) {
 void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_task) {
     ring_py::TensorMeta meta;
     if (!fifo_.pop(meta)) {
-        return;
+        throw std::runtime_error("P2PThread: tensor metadata queue is empty");
     }
 
     // Get step context -- pop from context queue if this is the first
     // hook in a new step (current_ctx_ is null).
     if (!current_ctx_) {
         current_ctx_ = fifo_.pop_context();
-        if (!current_ctx_) return;  // no context available
+        if (!current_ctx_) {
+            throw std::runtime_error("P2PThread: step context queue is empty");
+        }
     }
 
     if (meta.shape.empty() || first_task.tensor_total_bytes == 0) {
@@ -237,11 +214,11 @@ void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_ta
     }
 
     if (static_cast<uint64_t>(expected_bytes) != first_task.tensor_total_bytes) {
-        fprintf(stderr, "[p2p] WARN: shape/bytes mismatch: expected=%ld actual=%lu hook=%s\n",
-                (long)expected_bytes, (unsigned long)first_task.tensor_total_bytes,
-                ring_py::hook_type_name(meta.hook_type));
-        if (meta.last_in_step) { delete current_ctx_; current_ctx_ = nullptr; }
-        return;
+        throw std::runtime_error(
+            "P2PThread: shape/bytes mismatch for " +
+            std::string(ring_py::hook_type_name(meta.hook_type)) +
+            ": expected=" + std::to_string(expected_bytes) +
+            " actual=" + std::to_string(first_task.tensor_total_bytes));
     }
 
     tensor = tensor.view(dtype).reshape(
@@ -306,21 +283,10 @@ void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_ta
             slice = slice.clone();
         }
 
-        try {
-            submit_fn_(current_ctx_->model_id, shard_rank,
-                       req.req_id, act_name, meta.layer_no,
-                       db_start, db_end,
-                       std::move(slice));
-        } catch (const std::exception& e) {
-            log_submit_failure_once(current_ctx_->model_id, req.req_id,
-                                    act_name, meta.layer_no, shard_rank,
-                                    db_start, db_end, e.what());
-        } catch (...) {
-            log_submit_failure_once(current_ctx_->model_id, req.req_id,
-                                    act_name, meta.layer_no, shard_rank,
-                                    db_start, db_end,
-                                    "unknown non-std exception");
-        }
+        submit_fn_(current_ctx_->model_id, shard_rank,
+                   req.req_id, act_name, meta.layer_no,
+                   db_start, db_end,
+                   std::move(slice));
     }
 
     // Last hook in step -- free context
diff --git a/native/csrc/ring/ring_config.h b/native/csrc/ring/ring_config.h
index 6f02fc875..9e6d6d805 100644
--- a/native/csrc/ring/ring_config.h
+++ b/native/csrc/ring/ring_config.h
@@ -36,7 +36,7 @@ struct DrainFlushConfig {
 
     // Time-based flush: if a complete tensor has been pending for longer
     // than this many microseconds, flush unconditionally.  0 = disabled.
-    uint64_t timeout_us = 0;
+    uint64_t timeout_us = 100000;
 };
 
 // ---------------------------------------------------------------------------
@@ -65,8 +65,7 @@ struct RingConfig {
     // that keep the full tensor alive until consumed.
     bool clone_slices = false;
 
-    // ClickHouse insert queue limits (host engine).
-    // P2p thread blocks on submit_direct() when queue is full.
+    // Reserved for compatibility. StageConfig.input_queue owns host limits.
     uint64_t insert_queue_max_bytes = 4096ULL * 1024 * 1024;  // 4 GiB
     uint64_t insert_queue_max_items = 65536;
 
diff --git a/native/csrc/ring/ring_engine.cu b/native/csrc/ring/ring_engine.cu
index 2b26cf903..3bd876671 100644
--- a/native/csrc/ring/ring_engine.cu
+++ b/native/csrc/ring/ring_engine.cu
@@ -3,7 +3,9 @@
 
 #include "ring_engine.h"
 
+#include <exception>
 #include <stdexcept>
+#include <string>
 
 namespace ring {
 
@@ -48,20 +50,49 @@ void RingEngine::init(cudaStream_t stream) {
 }
 
 void RingEngine::start() {
+    if (started_.exchange(true)) return;
     drain_->start();
     p2p_->start();
 }
 
 void RingEngine::stop() {
-    // Guard against double-stop (benchmark _timed_close + engine.close).
-    if (!drain_->is_running()) return;
+    if (!started_.exchange(false)) {
+        drain_->rethrow_if_failed();
+        return;
+    }
 
-    cudaDeviceSynchronize();
-    drain_->force_flush_and_wait();
+    std::exception_ptr failure;
+    auto capture = [&failure] {
+        if (!failure) failure = std::current_exception();
+    };
 
+    if (!drain_->has_failed()) {
+        cudaError_t error = cudaDeviceSynchronize();
+        if (error != cudaSuccess) {
+            try {
+                throw std::runtime_error(
+                    std::string("RingEngine: cudaDeviceSynchronize failed: ") +
+                    cudaGetErrorString(error));
+            } catch (...) {
+                capture();
+            }
+        } else {
+            try {
+                drain_->force_flush_and_wait();
+            } catch (...) {
+                capture();
+            }
+        }
+    }
     drain_->stop();
     drain_->signal_p2p_stop();
     p2p_->stop();
+    try {
+        drain_->rethrow_if_failed();
+    } catch (...) {
+        capture();
+    }
+    if (failure) std::rethrow_exception(failure);
 }
 
 }  // namespace ring
diff --git a/native/csrc/ring/ring_engine.h b/native/csrc/ring/ring_engine.h
index 4a56a5f9a..64ce757d4 100644
--- a/native/csrc/ring/ring_engine.h
+++ b/native/csrc/ring/ring_engine.h
@@ -7,6 +7,7 @@
 #include "p2p_thread.h"
 #include "tensor_meta.h"
 
+#include <atomic>
 #include <memory>
 #include <vector>
 
@@ -38,6 +39,7 @@ class RingEngine {
     PinnedStaging   staging_;
     std::unique_ptr<DrainThread>  drain_;
     std::unique_ptr<P2PThread>    p2p_;
+    std::atomic<bool>             started_{false};
 };
 
 }  // namespace ring
diff --git a/native/csrc/ring/ring_engine_py.cu b/native/csrc/ring/ring_engine_py.cu
index c91c6d830..40998ba13 100644
--- a/native/csrc/ring/ring_engine_py.cu
+++ b/native/csrc/ring/ring_engine_py.cu
@@ -11,6 +11,8 @@
 #include "ring/producer.cuh"
 #include "ring/ring_debug.h"
 #include <ATen/cuda/CUDAContext.h>  // at::cuda::getCurrentCUDAStream
+#include <stdexcept>
+#include <string>
 
 // Forward-declare symbols from producer.cu
 namespace ring {
@@ -19,6 +21,14 @@ void set_ring_null_mode(bool enabled);
 
 namespace ring_py {
 
+static void check_cuda(cudaError_t error, const char* operation) {
+    if (error != cudaSuccess) {
+        throw std::runtime_error(
+            std::string("RingEngine: ") + operation + " failed: " +
+            cudaGetErrorString(error));
+    }
+}
+
 // ---------------------------------------------------------------------------
 struct RingEnginePy::Impl {
     TensorMetaFifo   fifo;
@@ -43,7 +53,7 @@ struct RingEnginePy::Impl {
     {
         const auto& state = engine.ring_state();
         int dev_idx = 0;
-        cudaGetDevice(&dev_idx);
+        check_cuda(cudaGetDevice(&dev_idx), "cudaGetDevice");
         payload_view = at::from_blob(
             state.payload_buf,
             {static_cast<int64_t>(state.payload_cap)},
@@ -97,9 +107,9 @@ void RingEnginePy::set_null_mode(bool enabled) {
     // NOT synchronize with PyTorch's non-blocking compute streams.  Sync
     // before to drain pending producer kernels that need the old value,
     // and after to ensure the new value is visible before the next launch.
-    cudaDeviceSynchronize();
+    check_cuda(cudaDeviceSynchronize(), "cudaDeviceSynchronize");
     ring::set_ring_null_mode(enabled);
-    cudaDeviceSynchronize();
+    check_cuda(cudaDeviceSynchronize(), "cudaDeviceSynchronize");
 }
 
 
@@ -174,8 +184,8 @@ void RingEnginePy::notify_drain() {
 // ---------------------------------------------------------------------------
 // prepare_step -- single Python->C++ call for pre-forward capacity check.
 //
-// Fast path (STEP_RING_OK): reads two uint64_t counters, returns immediately.
-// No stream resolution, no sync, no flush.
+// Fast path (STEP_RING_OK): reads one locked capacity snapshot.
+// No CUDA stream resolution, sync, or flush.
 //
 // Slow path (STEP_RING_FLUSHED / STEP_OVERSIZED): resolves the current CUDA
 // stream via at::cuda::getCurrentCUDAStream(), synchronises it, then asks the
@@ -224,28 +234,23 @@ int RingEnginePy::prepare_step(uint64_t step_total_bytes,
     // net starts firing.
     if (step_total_bytes > effective_cap || num_hooks > tcap) {
         cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-        cudaStreamSynchronize(ms);
+        check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
         drain.force_flush_and_wait();
         return STEP_OVERSIZED;
     }
 
-    // Case A: step fits.  Check available space for BOTH payload AND tasks.
-    const uint64_t payload_avail = pcap -
-        (drain.cpu_payload_head() - drain.cpu_payload_tail_committed());
-    const uint64_t task_avail = tcap -
-        (drain.cpu_task_head() - drain.cpu_task_tail_committed());
-
-    if (step_total_bytes <= payload_avail && num_hooks <= task_avail) {
-        drain.reserve(step_total_bytes, num_hooks);
-        return STEP_RING_OK;  // fast path -- no CUDA or thread interaction
+    if (drain.try_reserve(step_total_bytes, num_hooks, pcap, tcap)) {
+        return STEP_RING_OK;  // fast path -- no CUDA interaction
     }
 
     // Either payload or task ring full from prior steps.  Sync main
     // stream so all producer kernels finish writing, then flush.
     cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-    cudaStreamSynchronize(ms);
+    check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
     drain.force_flush_and_wait();
-    drain.reserve(step_total_bytes, num_hooks);
+    if (!drain.try_reserve(step_total_bytes, num_hooks, pcap, tcap)) {
+        throw std::logic_error("RingEngine: empty ring cannot fit step");
+    }
     return STEP_RING_FLUSHED;
 }
 
@@ -277,43 +282,16 @@ at::Tensor RingEnginePy::payload_tensor() const {
 // HookPoint.forward.  All three are called only when force_eager is active
 // (eager mode); never run during CUDA-graph capture or replay.
 //
-// Thread safety of the check-and-reserve pattern used by the safety net:
-//
-//   if nbytes <= available_capacity():
-//       reserve_one(nbytes)
-//
-// The main thread (this thread) is the only writer of cpu_payload_head_
-// (it advances only through reserve / reserve_one calls).  The drain
-// thread only ever advances cpu_payload_tail_committed_ forward as it
-// frees ring space.  Between the check and the reserve:
-//   - tail may move forward (drain freed more): actual available at
-//     reserve time is >= what we observed.
-//   - head is unchanged (single-threaded writer).
-// So the check's "fits" decision remains valid at reserve time.  No extra
-// locking around the pair is required.
-//
-// Within available_capacity(), the two accessor calls happen under
-// separate mutex acquires (drain.cpu_payload_head() and
-// drain.cpu_payload_tail_committed() each take mgmt_mu_ internally).
-// The observed snapshot is non-atomic: if drain advances tail between
-// the two reads, available_observed = pcap - head + tail_later, which
-// is >= the true available at the time of the head read.  That is, the
-// non-atomicity errs on the "over-estimate available" side -- the
-// reserve will still succeed because the actual ring state has at least
-// as much room as we computed.
+// try_reserve_one() checks payload, staging, and task capacity atomically.
 // ---------------------------------------------------------------------------
 
-uint64_t RingEnginePy::available_capacity() const {
-    auto& drain = impl_->engine.drain_thread();
-    const uint64_t pcap = impl_->engine.payload_cap();
-    return pcap - (drain.cpu_payload_head() - drain.cpu_payload_tail_committed());
+uint64_t RingEnginePy::effective_capacity() const {
+    return std::min(impl_->engine.payload_cap(), impl_->engine.staging_cap());
 }
 
-// Per-hook reservation: claim nbytes of payload + 1 task entry for an
-// upcoming producer kernel launch.  Caller must have checked
-// available_capacity() first.  drain.reserve takes mgmt_mu_ internally.
-void RingEnginePy::reserve_one(uint64_t nbytes) {
-    impl_->engine.drain_thread().reserve(nbytes, 1);
+bool RingEnginePy::try_reserve_one(uint64_t nbytes) {
+    return impl_->engine.drain_thread().try_reserve(
+        nbytes, 1, effective_capacity(), impl_->engine.task_cap());
 }
 
 // Synchronise the current CUDA stream so all queued producer kernels
@@ -322,7 +300,7 @@ void RingEnginePy::reserve_one(uint64_t nbytes) {
 // Python binding releases the GIL.
 void RingEnginePy::flush_and_wait() {
     cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-    cudaStreamSynchronize(ms);
+    check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
     impl_->engine.drain_thread().force_flush_and_wait();
 }
 
diff --git a/native/csrc/ring/ring_engine_py.h b/native/csrc/ring/ring_engine_py.h
index e96e1e4a6..9f10f1ceb 100644
--- a/native/csrc/ring/ring_engine_py.h
+++ b/native/csrc/ring/ring_engine_py.h
@@ -30,10 +30,10 @@ struct RingConfig {
     float    drain_flush_payload_ratio  = 0.5f;
     uint64_t drain_flush_entry_threshold = 0;
     uint64_t drain_flush_byte_threshold  = 0;
-    uint64_t drain_flush_timeout_us      = 0;
+    uint64_t drain_flush_timeout_us      = 100000;
     // Clone per-request slices
     bool     clone_slices               = false;
-    // ClickHouse insert queue limits
+    // Reserved for compatibility; StageConfig owns host queue limits.
     uint64_t insert_queue_max_bytes     = 4096ULL * 1024 * 1024;
     uint64_t insert_queue_max_items     = 65536;
 };
@@ -149,15 +149,10 @@ class RingEnginePy {
     //      HookPoint.forward (eager-only path).  Never called during
     //      CUDA-graph capture or replay.
 
-    // Free bytes in the payload ring not currently reserved and not
-    // pending drain.  CPU-only read.
-    uint64_t available_capacity() const;
+    uint64_t effective_capacity() const;
 
-    // Per-hook reservation: claim `nbytes` of payload ring + 1 task entry
-    // for an upcoming producer kernel launch.  Used by the safety net
-    // when force_eager is on and the spec is dynamic-shape.  Advances
-    // cpu_payload_head/cpu_task_head atomically.
-    void reserve_one(uint64_t nbytes);
+    // Atomically check and reserve payload plus one task slot.
+    bool try_reserve_one(uint64_t nbytes);
 
     // Synchronise the current CUDA stream + force drain to process all
     // outstanding entries.  Blocking; the Python binding releases the
diff --git a/native/csrc/ring/task_entry.h b/native/csrc/ring/task_entry.h
index f56ab0992..73199b543 100644
--- a/native/csrc/ring/task_entry.h
+++ b/native/csrc/ring/task_entry.h
@@ -17,7 +17,7 @@ namespace ring {
 // Sentinel value for ready_seq -- indicates slot has not been published yet.
 //
 // Publish protocol:
-//   producer: write all data fields -> __threadfence() -> write ready_seq = seq_no
+//   producer: write fields -> __threadfence_system() -> publish ready_seq
 //   consumer: poll until __atomic_load_n(ready_seq) == expected -> read fields
 // ---------------------------------------------------------------------------
 static constexpr uint64_t READY_SEQ_SENTINEL = ~uint64_t(0);
diff --git a/native/csrc/ring/task_ring.cuh b/native/csrc/ring/task_ring.cuh
index 9e6978817..4897eea05 100644
--- a/native/csrc/ring/task_ring.cuh
+++ b/native/csrc/ring/task_ring.cuh
@@ -12,7 +12,7 @@
 //
 // Publish protocol (producer):
 //   1. Write all TaskEntry data fields at slot (head % capacity).
-//   2. __threadfence() -- ensures data is visible before ready_seq.
+//   2. __threadfence_system() -- makes data visible to the CPU consumer.
 //   3. Write ready_seq = head  (the slot's logical sequence number).
 //   4. Increment head.
 //
@@ -73,7 +73,7 @@ inline void task_ring_init(TaskEntry* d_entries, uint64_t capacity,
 // task_publish -- write a TaskEntry and publish it to the consumer.
 //
 // Copies all non-ready_seq fields from `src` into the slot at `seq_no %
-// capacity`, issues a __threadfence() to enforce write ordering, then writes
+// capacity`, issues a system-scope fence to enforce write ordering, then writes
 // ready_seq = seq_no.
 // ---------------------------------------------------------------------------
 __device__ inline void task_publish(
@@ -93,7 +93,7 @@ __device__ inline void task_publish(
     slot.payload_len2       = src.payload_len2;
 
     // Release fence: all stores above must be visible before ready_seq.
-    __threadfence();
+    __threadfence_system();
 
     // Publish: consumer spins until it sees this value.
     *reinterpret_cast<volatile uint64_t*>(&slot.ready_seq) = seq_no;
diff --git a/src/dmi/engine.py b/src/dmi/engine.py
index 0405063dd..42aa90d87 100644
--- a/src/dmi/engine.py
+++ b/src/dmi/engine.py
@@ -9,7 +9,7 @@
 from .config import MonitoringConfig
 
 
-DEFAULT_DRAIN_FLUSH_TIMEOUT_US = 0
+DEFAULT_DRAIN_FLUSH_TIMEOUT_US = 100_000
 
 
 def _native_module() -> Any:
@@ -280,6 +280,13 @@ def next_auto_group_id(self) -> int:
     def close(self) -> None:
         """Tear down backend resources."""
 
+        first_error: Optional[BaseException] = None
+
+        def capture_error(exc: BaseException) -> None:
+            nonlocal first_error
+            if first_error is None:
+                first_error = exc
+
         if self._ring_transport is not None:
             # Best-effort reset of the device-global native null flag.  This is
             # needed only after callers explicitly disabled capture; the normal
@@ -287,30 +294,43 @@ def close(self) -> None:
             if not self.capture_enabled:
                 try:
                     self.set_capture_enabled(True)
-                except Exception:
-                    pass
+                except Exception as exc:
+                    capture_error(exc)
             try:
                 ring_engine = getattr(self, "_ring_engine", None)
                 if ring_engine is not None:
                     ring_engine.stop()
-            except Exception:
-                pass
+            except Exception as exc:
+                capture_error(exc)
             try:
                 _rt = _ring_module()
                 _rt.deactivate()
-            except Exception:
-                pass
+            except Exception as exc:
+                capture_error(exc)
             self._ring_transport = None
             self._ring_engine = None
 
         if self._host_engine is not None:
+            host_engine = self._host_engine
             try:
-                self._host_engine.close_input()
-                self._host_engine.stop()
-            except Exception:
-                pass
+                host_engine.close_input()
+            except Exception as exc:
+                capture_error(exc)
+            try:
+                host_engine.stop()
+            except Exception as exc:
+                capture_error(exc)
+            try:
+                raise_if_failed = getattr(host_engine, "raise_if_failed", None)
+                if raise_if_failed is not None:
+                    raise_if_failed()
+            except Exception as exc:
+                capture_error(exc)
             self._host_engine = None
 
+        if first_error is not None:
+            raise first_error
+
 
 # ---------------------------------------------------------------------------
 # Backend loader
diff --git a/src/dmi/hooks/point.py b/src/dmi/hooks/point.py
index e5c42a122..8c25cd376 100644
--- a/src/dmi/hooks/point.py
+++ b/src/dmi/hooks/point.py
@@ -283,13 +283,13 @@ def forward(self, x: Tensor) -> Tensor:
             engine = transport._ring_engine
             if engine is not None:
                 nbytes = x_cont.nbytes
-                if nbytes <= engine.available_capacity():
-                    engine.reserve_one(nbytes)
+                if engine.try_reserve_one(nbytes):
                     dispatch_producer(ring_payload, x_cont, strip_t, strip_rb,
                                       self._ring_hook_type, self._ring_hook_id)
-                elif nbytes <= engine.payload_cap():
+                elif nbytes <= engine.effective_capacity():
                     engine.flush_and_wait()
-                    engine.reserve_one(nbytes)
+                    if not engine.try_reserve_one(nbytes):
+                        raise RuntimeError("Ring reservation failed after flush")
                     dispatch_producer(ring_payload, x_cont, strip_t, strip_rb,
                                       self._ring_hook_type, self._ring_hook_id)
                 else:
diff --git a/src/dmi/transport/ring.py b/src/dmi/transport/ring.py
index 7b276a0ec..44ddadc14 100644
--- a/src/dmi/transport/ring.py
+++ b/src/dmi/transport/ring.py
@@ -164,8 +164,8 @@ def __init__(self, ring_engine: Any) -> None:
 
         # When True, HookPoint.forward takes the runtime safety-net branch
         # instead of the fast path:
-        #   1. fits in current slack       -> reserve_one + ring
-        #   2. fits after flushing the ring -> flush_and_wait + reserve_one + ring
+        #   1. fits in current slack        -> try_reserve_one + ring
+        #   2. fits after flushing the ring -> flush + try_reserve_one + ring
         #   3. single tensor > ring        -> flush_and_wait + submit_cpu_direct
         # Owned by adaptor_base.before_forward (per-batch reassignment based
         # on prepare_step result and dynamic-spec presence).  Dispatch
diff --git a/tests/native/ring/Makefile b/tests/native/ring/Makefile
index 1a1b068b0..2c2b863f9 100644
--- a/tests/native/ring/Makefile
+++ b/tests/native/ring/Makefile
@@ -54,6 +54,7 @@ NVCC_FLAGS := \
 
 PRODUCER_CU  := ../../../native/csrc/ring/producer.cu
 DRAIN_SRC    := ../../../native/csrc/ring/drain_thread.cpp
+P2P_SRC      := ../../../native/csrc/ring/p2p_thread.cpp
 
 TARGETS := $(BUILD)/test_rings $(BUILD)/test_producer \
            $(BUILD)/test_ring_engine $(BUILD)/test_null_mode
@@ -76,7 +77,7 @@ $(BUILD)/test_rings: test_rings.cu $(CUDA_CONFIG_STAMP) | $(BUILD)
 $(BUILD)/test_producer: test_producer.cu $(PRODUCER_CU) $(CUDA_CONFIG_STAMP) | $(BUILD)
 	$(CUDA_NVCC) $(NVCC_FLAGS) $(filter-out $(CUDA_CONFIG_STAMP),$^) -o $@
 
-$(BUILD)/test_ring_engine: test_ring_engine.cu $(PRODUCER_CU) $(DRAIN_SRC) $(CUDA_CONFIG_STAMP) | $(BUILD)
+$(BUILD)/test_ring_engine: test_ring_engine.cu $(PRODUCER_CU) $(DRAIN_SRC) $(P2P_SRC) $(CUDA_CONFIG_STAMP) | $(BUILD)
 	$(CUDA_NVCC) $(NVCC_FLAGS) $(TORCH_INCLUDE_FLAGS) $(TORCH_LIB_FLAGS) \
 		$(TORCH_RPATH_FLAGS) -ltorch -ltorch_cpu -lc10 -lpthread \
 		$(filter-out $(CUDA_CONFIG_STAMP),$^) -o $@
diff --git a/tests/native/ring/test_producer.cu b/tests/native/ring/test_producer.cu
index 59012d78e..df9056913 100644
--- a/tests/native/ring/test_producer.cu
+++ b/tests/native/ring/test_producer.cu
@@ -6,6 +6,7 @@
 #include <cuda_runtime.h>
 
 #include <algorithm>
+#include <chrono>
 #include <cstdint>
 #include <cstdio>
 #include <cstdlib>
@@ -276,6 +277,31 @@ static void test_serialized_static_launches() {
     }
 }
 
+static void test_host_observes_complete_published_descriptor() {
+    banner("host observes complete published descriptor without device sync");
+    ring::AllocatedRing allocated(make_config());
+    allocated.init();
+    ring::RingState& state = allocated.state();
+    const std::vector<uint8_t> source = pattern(333, 91);
+    uint8_t* device = upload(source);
+    cudaStream_t stream{};
+    CUDA_CHECK(cudaStreamCreateWithFlags(&stream, cudaStreamNonBlocking));
+
+    ring::launch_producer_static(state, device, source.size(), 5, stream);
+    const auto deadline = std::chrono::steady_clock::now() +
+                          std::chrono::seconds(5);
+    while (!ring::task_cpu_ready(state.task_entries, state.task_cap, 0) &&
+           std::chrono::steady_clock::now() < deadline) {
+    }
+
+    EXPECT(ring::task_cpu_ready(state.task_entries, state.task_cap, 0));
+    expect_entry(state, 0, source.size());
+
+    CUDA_CHECK(cudaStreamSynchronize(stream));
+    CUDA_CHECK(cudaStreamDestroy(stream));
+    CUDA_CHECK(cudaFree(device));
+}
+
 int main() {
     setbuf(stdout, nullptr);
     ring::set_ring_null_mode(false);
@@ -288,6 +314,7 @@ int main() {
     test_prefix_bounds();
     test_chunked_packed_copy();
     test_serialized_static_launches();
+    test_host_observes_complete_published_descriptor();
 
     std::printf("Results: %d passed, %d failed\n", g_pass, g_fail);
     return g_fail == 0 ? 0 : 1;
diff --git a/tests/native/ring/test_ring_engine.cu b/tests/native/ring/test_ring_engine.cu
index 806a85485..ee2535d27 100644
--- a/tests/native/ring/test_ring_engine.cu
+++ b/tests/native/ring/test_ring_engine.cu
@@ -1,17 +1,22 @@
 // CUDA integration tests for producer -> drain -> pinned-staging delivery.
 
 #include "ring/drain_thread.h"
+#include "ring/p2p_thread.h"
 #include "ring/pinned_staging.h"
 #include "ring/producer.cuh"
 #include "ring/ring_alloc.h"
 
 #include <cuda_runtime.h>
 
+#include <atomic>
+#include <chrono>
 #include <cstdint>
 #include <cstdio>
 #include <cstdlib>
 #include <cstring>
 #include <memory>
+#include <stdexcept>
+#include <thread>
 #include <vector>
 
 static int g_pass = 0;
@@ -148,10 +153,11 @@ static void test_static_force_flush() {
     EXPECT(task.alloc_bytes == reserved);
     EXPECT(task.data_len1 + task.data_len2 == source.size());
     EXPECT(task_bytes(task) == source);
-    EXPECT(harness.drain->cpu_task_head() == 1);
-    EXPECT(harness.drain->cpu_task_tail_committed() == 1);
-    EXPECT(harness.drain->cpu_payload_head() == reserved);
-    EXPECT(harness.drain->cpu_payload_tail_committed() == reserved);
+    const auto capacity = harness.drain->capacity_snapshot();
+    EXPECT(capacity.task_head == 1);
+    EXPECT(capacity.task_tail_committed == 1);
+    EXPECT(capacity.payload_head == reserved);
+    EXPECT(capacity.payload_tail_committed == reserved);
 
     harness.release(task);
     CUDA_CHECK(cudaFree(device));
@@ -180,7 +186,7 @@ static void test_prefix_force_flush() {
     EXPECT(task.tensor_total_bytes == actual);
     EXPECT(task.alloc_bytes == actual);
     EXPECT(task_bytes(task) == expected);
-    EXPECT(harness.drain->cpu_payload_tail_committed() == actual);
+    EXPECT(harness.drain->capacity_snapshot().payload_tail_committed == actual);
 
     harness.release(task);
     CUDA_CHECK(cudaFree(device_count));
@@ -214,8 +220,9 @@ static void test_repeated_wrap_delivery() {
 
     EXPECT(*harness.allocated.state().task_head == 4);
     EXPECT(*harness.allocated.state().payload_head == 320);
-    EXPECT(harness.drain->cpu_task_tail_committed() == 4);
-    EXPECT(harness.drain->cpu_payload_tail_committed() == 320);
+    const auto capacity_snapshot = harness.drain->capacity_snapshot();
+    EXPECT(capacity_snapshot.task_tail_committed == 4);
+    EXPECT(capacity_snapshot.payload_tail_committed == 320);
     EXPECT(harness.staging.head() == 320);
     EXPECT(harness.staging.tail() == 320);
 }
@@ -234,11 +241,123 @@ static void test_zero_byte_delivery() {
     EXPECT(task.data_len1 == 0);
     EXPECT(task.data_ptr2 == nullptr);
     EXPECT(task.data_len2 == 0);
-    EXPECT(harness.drain->cpu_task_tail_committed() == 1);
-    EXPECT(harness.drain->cpu_payload_tail_committed() == 0);
+    const auto capacity = harness.drain->capacity_snapshot();
+    EXPECT(capacity.task_tail_committed == 1);
+    EXPECT(capacity.payload_tail_committed == 0);
     harness.release(task);
 }
 
+static void test_capacity_snapshot_is_consistent_during_reserve() {
+    banner("capacity snapshot is consistent during reserve");
+    DrainHarness harness(make_config(1ULL << 24));
+    constexpr uint64_t reservations = 10000;
+    constexpr uint64_t bytes_per_reservation = 64;
+    std::atomic<bool> done{false};
+
+    std::thread writer([&] {
+        for (uint64_t i = 0; i < reservations; ++i) {
+            harness.drain->reserve(bytes_per_reservation, 1);
+        }
+        done.store(true, std::memory_order_release);
+    });
+
+    while (!done.load(std::memory_order_acquire)) {
+        const auto snapshot = harness.drain->capacity_snapshot();
+        EXPECT(snapshot.payload_head ==
+               snapshot.task_head * bytes_per_reservation);
+    }
+    writer.join();
+
+    const auto snapshot = harness.drain->capacity_snapshot();
+    EXPECT(snapshot.payload_head == reservations * bytes_per_reservation);
+    EXPECT(snapshot.task_head == reservations);
+    EXPECT(snapshot.payload_tail_committed == 0);
+    EXPECT(snapshot.task_tail_committed == 0);
+}
+
+static void test_atomic_reservation_checks_effective_and_task_capacity() {
+    banner("atomic reservation checks effective and task capacity");
+    DrainHarness harness(make_config());
+
+    EXPECT(!harness.drain->try_reserve(257, 1, 256, 16));
+    for (uint32_t i = 0; i < 16; ++i) {
+        EXPECT(harness.drain->try_reserve(0, 1, 256, 16));
+    }
+    EXPECT(!harness.drain->try_reserve(0, 1, 256, 16));
+}
+
+static void test_force_flush_surfaces_invalid_d2h_source() {
+    banner("force flush surfaces invalid D2H source");
+    DrainHarness harness(make_config());
+    ring::RingState& state = harness.allocated.state();
+    uint8_t* payload_buf = state.payload_buf;
+    state.payload_buf = reinterpret_cast<uint8_t*>(1);
+
+    ring::TaskEntry& entry = state.task_entries[0];
+    entry.tensor_total_bytes = 64;
+    entry.payload_off1 = 0;
+    entry.payload_len1 = 64;
+    entry.payload_off2 = 0;
+    entry.payload_len2 = 0;
+    __atomic_store_n(&entry.ready_seq, uint64_t{0}, __ATOMIC_RELEASE);
+    harness.drain->reserve(64, 1);
+
+    bool threw = false;
+    try {
+        harness.drain->force_flush_and_wait();
+    } catch (const std::runtime_error&) {
+        threw = true;
+    }
+    EXPECT(threw);
+
+    state.payload_buf = payload_buf;
+}
+
+static void throwing_submit(
+    const std::string&, int32_t, const std::string&, const std::string&,
+    int32_t, int32_t, int32_t, at::Tensor)
+{
+    throw std::runtime_error("host submission failed");
+}
+
+static void test_p2p_submission_failure_reaches_drain_owner() {
+    banner("P2P submission failure reaches drain owner");
+    DrainHarness harness(make_config());
+    ring_py::TensorMetaFifo fifo;
+    auto* context = new ring_py::StepContext();
+    context->model_id = "test";
+    context->requests.push_back({"request", 0, 1, 0, 0});
+    std::vector<ring_py::TensorMeta> metas(1);
+    metas[0].hook_type = ring_py::HOOK_TYPE_TOKEN_IDS;
+    metas[0].shape = {1, 1};
+    metas[0].dtype = static_cast<int>(at::kByte);
+    metas[0].last_in_step = true;
+    fifo.push_step(context, metas);
+
+    ring::P2PThread p2p(*harness.drain, fifo, harness.cfg, throwing_submit);
+    p2p.start();
+    harness.drain->submit_cpu_direct(
+        at::zeros({1, 1}, at::TensorOptions().dtype(at::kByte)), 1);
+
+    const auto deadline = std::chrono::steady_clock::now() +
+                          std::chrono::seconds(5);
+    while (!harness.drain->has_failed() &&
+           std::chrono::steady_clock::now() < deadline) {
+        std::this_thread::yield();
+    }
+    EXPECT(harness.drain->has_failed());
+    bool threw = false;
+    try {
+        harness.drain->rethrow_if_failed();
+    } catch (const std::runtime_error&) {
+        threw = true;
+    }
+    EXPECT(threw);
+
+    harness.drain->signal_p2p_stop();
+    p2p.stop();
+}
+
 int main() {
     setbuf(stdout, nullptr);
     ring::set_ring_null_mode(false);
@@ -249,6 +368,10 @@ int main() {
     test_prefix_force_flush();
     test_repeated_wrap_delivery();
     test_zero_byte_delivery();
+    test_capacity_snapshot_is_consistent_during_reserve();
+    test_atomic_reservation_checks_effective_and_task_capacity();
+    test_force_flush_surfaces_invalid_d2h_source();
+    test_p2p_submission_failure_reaches_drain_owner();
 
     std::printf("Results: %d passed, %d failed\n", g_pass, g_fail);
     return g_fail == 0 ? 0 : 1;
diff --git a/tests/test_clickhouse_host_benchmark.py b/tests/test_clickhouse_host_benchmark.py
index 7e3ed77be..06407c437 100644
--- a/tests/test_clickhouse_host_benchmark.py
+++ b/tests/test_clickhouse_host_benchmark.py
@@ -9,6 +9,7 @@
 import benchmarks.bench_clickhouse_host as benchmark
 from benchmarks.bench_clickhouse_host import (
     BenchmarkConfig,
+    ServerTelemetrySampler,
     TrialMeasurement,
     _ensure_table,
     _parts_metrics,
@@ -18,8 +19,11 @@
     build_client_settings,
     configure_stage,
     generate_payload_pool,
+    parse_parallelism_sweep,
     parse_byte_size,
     quote_identifier,
+    run_sweep,
+    summarize_scaling_trials,
 )
 
 pytestmark = pytest.mark.cpu
@@ -78,6 +82,13 @@ def test_config_rejects_non_positive_index_granularity():
         BenchmarkConfig(index_granularity=0)
 
 
+def test_config_rejects_invalid_socket_and_sampling_timeouts():
+    with pytest.raises(ValueError, match="socket_timeout_seconds"):
+        BenchmarkConfig(socket_timeout_seconds=0)
+    with pytest.raises(ValueError, match="server_sample_interval_ms"):
+        BenchmarkConfig(server_sample_interval_ms=-1)
+
+
 def test_config_uses_a_unique_benchmark_table_by_default():
     first = BenchmarkConfig().table
     second = BenchmarkConfig().table
@@ -132,16 +143,31 @@ def test_stage_configuration_sets_batching_and_backpressure():
     assert queue.high_watermark_items == 100
     assert queue.high_watermark_size == 256
     assert policy.block is True
+    assert policy.timeout_s == cfg.socket_timeout_seconds
 
 
 def test_async_insert_settings_wait_for_durability():
-    assert build_client_settings(False) == {}
+    assert build_client_settings(False) == {"async_insert": 0}
     assert build_client_settings(True) == {
         "async_insert": 1,
         "wait_for_async_insert": 1,
     }
 
 
+def test_single_value_sweep_uses_requested_parallelism(monkeypatch, capsys):
+    seen = []
+
+    def runner(config):
+        seen.append(config.parallelism)
+        return {"parallelism": config.parallelism}
+
+    monkeypatch.setattr(benchmark, "run", runner)
+
+    assert benchmark.main(["--parallelism-sweep", "8"]) == 0
+    assert seen == [8]
+    assert '"parallelism": 8' in capsys.readouterr().out
+
+
 def test_server_time_uses_clickhouse_clock():
     class Client:
         def execute(self, query):
@@ -279,6 +305,8 @@ def test_trial_measurement_reports_enqueue_and_drain_separately():
         enqueue_seconds=2.0,
         total_seconds=5.0,
         enqueue_latencies_ns=(1_000_000, 2_000_000, 3_000_000, 4_000_000),
+        startup_seconds=0.25,
+        client_metrics={"peak_active_inserts": 3},
     )
 
     report = trial.as_dict()
@@ -289,6 +317,8 @@ def test_trial_measurement_reports_enqueue_and_drain_separately():
     assert report["enqueue"]["latency_ms"]["p50"] == 2.5
     assert report["enqueue"]["latency_ms"]["p95"] == pytest.approx(3.85)
     assert report["enqueue"]["latency_samples"] == 4
+    assert report["startup_seconds"] == 0.25
+    assert report["client"]["peak_active_inserts"] == 3
 
 
 def test_submit_and_drain_uses_graceful_lifecycle():
@@ -299,6 +329,10 @@ def __init__(self):
         def start(self):
             self.calls.append("start")
 
+        def wait_until_ready(self, timeout):
+            self.calls.append(("wait_until_ready", timeout))
+            return True
+
         def submit_direct(self, *args):
             self.calls.append(("submit", args))
 
@@ -312,10 +346,30 @@ def join(self, timeout):
         def raise_if_failed(self):
             self.calls.append("raise_if_failed")
 
+        def clickhouse_metrics(self):
+            return SimpleNamespace(
+                expected_workers=2,
+                ready_workers=2,
+                active_inserts=0,
+                peak_active_inserts=2,
+                batches=2,
+                rows=2,
+                logical_bytes=32,
+                insert_seconds=0.5,
+                workers=[],
+            )
+
         def request_abort(self):
             self.calls.append("request_abort")
 
-    ticks = iter((0, 10, 15, 20, 30, 40, 70))
+    class Sampler:
+        def start(self):
+            engine.calls.append("sampler_start")
+
+        def stop(self):
+            engine.calls.append("sampler_stop")
+
+    ticks = iter((0, 10, 20, 30, 40, 50, 60, 70, 90))
     engine = Engine()
     payload = SimpleNamespace(nbytes=16)
 
@@ -326,9 +380,12 @@ def request_abort(self):
         model_id="run",
         timeout_seconds=5.0,
         clock=lambda: next(ticks),
+        deadline_clock=lambda: 0.0,
+        sampler=Sampler(),
     )
 
     assert [call[0] for call in engine.calls if isinstance(call, tuple)] == [
+        "wait_until_ready",
         "submit",
         "submit",
         "join",
@@ -336,8 +393,144 @@ def request_abort(self):
     assert "close_input" in engine.calls
     assert "raise_if_failed" in engine.calls
     assert "request_abort" not in engine.calls
-    assert result.enqueue_seconds == 40 / 1_000_000_000
+    assert engine.calls.index("sampler_start") > engine.calls.index(("wait_until_ready", 5.0))
+    assert engine.calls.index("sampler_stop") > engine.calls.index(("join", 5.0))
+    assert result.startup_seconds == 10 / 1_000_000_000
+    assert result.enqueue_seconds == 50 / 1_000_000_000
     assert result.total_seconds == 70 / 1_000_000_000
+    assert result.client_metrics["peak_active_inserts"] == 2
+
+
+def test_submit_and_drain_does_not_restart_timeout_during_abort():
+    class Engine:
+        def __init__(self):
+            self.join_timeouts = []
+
+        def start(self):
+            pass
+
+        def wait_until_ready(self, timeout):
+            return True
+
+        def submit_direct(self, *args):
+            pass
+
+        def close_input(self):
+            pass
+
+        def join(self, timeout):
+            self.join_timeouts.append(timeout)
+            return False
+
+        def request_abort(self):
+            pass
+
+    deadlines = iter((0.0, 0.0, 0.0, 0.0, 5.0))
+    engine = Engine()
+
+    with pytest.raises(TimeoutError, match="drain exceeded"):
+        _submit_and_drain(
+            engine,
+            [SimpleNamespace(nbytes=16)],
+            rows=1,
+            model_id="run",
+            timeout_seconds=5.0,
+            clock=iter(range(20)).__next__,
+            deadline_clock=lambda: next(deadlines),
+        )
+
+    assert engine.join_timeouts == [5.0, 0.0]
+
+
+def test_parallelism_sweep_is_unique_and_positive():
+    assert parse_parallelism_sweep("1,2,4,2,8") == (1, 2, 4, 8)
+    with pytest.raises(ValueError, match="positive"):
+        parse_parallelism_sweep("1,0,4")
+
+
+def test_scaling_summary_reports_speedup_and_variance():
+    trials = [
+        {"parallelism": 1, "throughput": 1.0},
+        {"parallelism": 1, "throughput": 1.2},
+        {"parallelism": 2, "throughput": 1.8},
+        {"parallelism": 2, "throughput": 2.0},
+    ]
+
+    summary = summarize_scaling_trials(trials, plateau_threshold_percent=5.0)
+
+    assert summary[0]["parallelism"] == 1
+    assert summary[0]["median_gib_per_second"] == pytest.approx(1.1)
+    assert summary[1]["speedup_vs_one"] == pytest.approx(1.9 / 1.1)
+    assert summary[1]["gain_vs_previous_percent"] == pytest.approx((1.9 / 1.1 - 1) * 100)
+    assert summary[1]["plateau"] is False
+
+
+def test_scaling_summary_does_not_invent_one_worker_baseline():
+    summary = summarize_scaling_trials(
+        [
+            {"parallelism": 2, "throughput": 2.0},
+            {"parallelism": 4, "throughput": 3.0},
+        ]
+    )
+
+    assert summary[0]["baseline_parallelism"] == 2
+    assert summary[0]["speedup_vs_one"] is None
+    assert summary[1]["speedup_vs_baseline"] == pytest.approx(1.5)
+
+
+def test_run_sweep_keeps_raw_trials_and_uses_unique_tables():
+    calls = []
+
+    def runner(config):
+        calls.append(config)
+        return {
+            "measurement": {
+                "total": {"logical_gib_per_second": float(config.parallelism)}
+            }
+        }
+
+    report = run_sweep(
+        BenchmarkConfig(rows=1, payload_bytes=16, min_batch_bytes=16,
+                        max_batch_bytes=16, queue_capacity_bytes=32,
+                        warmup_rows=0),
+        parallelisms=(1, 2),
+        trials=2,
+        runner=runner,
+    )
+
+    assert len(report["trials"]) == 4
+    assert {trial["parallelism"] for trial in report["trials"]} == {1, 2}
+    assert len({config.table for config in calls}) == 4
+    assert report["summary"][1]["speedup_vs_one"] == 2.0
+
+
+def test_server_sampler_tracks_insert_concurrency_and_metric_peaks():
+    class Client:
+        def __init__(self):
+            self.calls = 0
+
+        def execute(self, query):
+            self.calls += 1
+            if "system.processes" in query:
+                return [(2,)]
+            if "system.asynchronous_metrics" in query:
+                return [("OSUserTimeNormalized", 0.5)]
+            return [("Merge", 3), ("Query", 5)]
+
+        def disconnect(self):
+            pass
+
+    sampler = ServerTelemetrySampler(lambda: Client(), interval_ms=1)
+    sampler.sample_once()
+    snapshot = sampler.snapshot()
+
+    assert snapshot["samples"] == 1
+    assert snapshot["peak_active_inserts"] == 2
+    assert snapshot["metric_peaks"] == {
+        "Merge": 3.0,
+        "Query": 5.0,
+        "async.OSUserTimeNormalized": 0.5,
+    }
 
 
 def test_identifier_quoting_rejects_sql_fragments():
diff --git a/tests/test_cpu_native_build.py b/tests/test_cpu_native_build.py
index d98363285..f99abf60f 100644
--- a/tests/test_cpu_native_build.py
+++ b/tests/test_cpu_native_build.py
@@ -50,6 +50,49 @@ def load_named(name):
     assert calls == ["_native_backend", "_host_backend"]
 
 
+def test_clickhouse_stage_has_bounded_batching_defaults():
+    from dmi.transport.native import ClickHouseClientConfig, StageConfig
+
+    stage = StageConfig.clickhouse_insert(ClickHouseClientConfig())
+    queue = stage.input_queue
+
+    assert queue.min_batch_items is None
+    assert queue.min_batch_size == 16 * 1024**2
+    assert queue.max_linger_s == pytest.approx(0.05)
+    assert queue.max_batch_items == 10_000
+    assert queue.max_batch_size is None
+    assert queue.high_watermark_items == 20_000
+    assert queue.high_watermark_size == 512 * 1024**2
+
+
+def test_clickhouse_client_exposes_socket_timeouts_and_worker_metrics():
+    from dmi.transport.native import ClickHouseClientConfig, DMXHostEngine, StageConfig
+
+    config = ClickHouseClientConfig()
+    assert config.connect_timeout_ms == 5000
+    assert config.receive_timeout_ms == 0
+    assert config.send_timeout_ms == 0
+
+    engine = DMXHostEngine(StageConfig.clickhouse_insert(config, parallelism=3))
+    metrics = engine.clickhouse_metrics()
+    assert metrics.expected_workers == 3
+    assert metrics.ready_workers == 0
+    assert metrics.peak_active_inserts == 0
+    assert [worker.worker_index for worker in metrics.workers] == [0, 1, 2]
+
+
+def test_engine_metrics_follow_mutated_stage_parallelism():
+    from dmi.transport.native import ClickHouseClientConfig, DMXHostEngine, StageConfig
+
+    stage = StageConfig.clickhouse_insert(ClickHouseClientConfig(), parallelism=1)
+    stage.parallelism = 3
+
+    metrics = DMXHostEngine(stage).clickhouse_metrics()
+
+    assert metrics.expected_workers == 3
+    assert [worker.worker_index for worker in metrics.workers] == [0, 1, 2]
+
+
 def test_ring_export_requires_full_backend(monkeypatch):
     from dmi.transport import native
 
diff --git a/tests/test_engine_runtime_api.py b/tests/test_engine_runtime_api.py
index 108fe7cf1..ff7e72e91 100644
--- a/tests/test_engine_runtime_api.py
+++ b/tests/test_engine_runtime_api.py
@@ -10,15 +10,17 @@
 
 import pytest
 
+import dmi.engine as engine_module
 from dmi.engine import MonitoringEngine, RingCapacities
 
 pytestmark = pytest.mark.cpu
 
 
 class _FakeRingEngine:
-    def __init__(self, transport=None, *, fail_null_mode=False):
+    def __init__(self, transport=None, *, fail_null_mode=False, fail_stop=False):
         self.transport = transport
         self.fail_null_mode = fail_null_mode
+        self.fail_stop = fail_stop
         self.null_mode_calls = []
         self.stop_calls = 0
         self.init_calls = 0
@@ -55,10 +57,12 @@ def start(self):
 
     def stop(self):
         self.stop_calls += 1
+        if self.fail_stop:
+            raise RuntimeError("ring drain failed")
 
 
 def _engine_with_fake_ring(
-    *, null_offload=False, force_eager=False, fail_null_mode=False
+    *, null_offload=False, force_eager=False, fail_null_mode=False, fail_stop=False
 ):
     engine = MonitoringEngine(enable_ring_transport=False)
     transport = SimpleNamespace(
@@ -68,6 +72,7 @@ def _engine_with_fake_ring(
     ring_engine = _FakeRingEngine(
         transport,
         fail_null_mode=fail_null_mode,
+        fail_stop=fail_stop,
     )
     engine._ring_transport = transport
     engine._ring_engine = ring_engine
@@ -90,6 +95,25 @@ def test_ring_capacities_is_frozen_snapshot_with_effective_limit():
         capacities.payload_bytes = 1
 
 
+def test_default_ring_flushes_small_workloads_with_bounded_linger(monkeypatch):
+    class _RingConfig:
+        pass
+
+    monkeypatch.setattr(
+        engine_module,
+        "_native_module",
+        lambda: SimpleNamespace(RingConfig=_RingConfig),
+    )
+
+    config = MonitoringEngine._make_default_ring_config(
+        payload_mb=64,
+        pinned_mb=32,
+        task_entries=128,
+    )
+
+    assert config.drain_flush_timeout_us == 100_000
+
+
 def test_capture_toggle_changes_metadata_flag_after_native_transition():
     engine, transport, ring_engine = _engine_with_fake_ring(force_eager=True)
     assert engine.capture_enabled is True
@@ -143,6 +167,75 @@ def test_close_restores_device_global_null_mode_before_ring_stop():
     assert engine.capture_enabled is False
 
 
+def test_close_surfaces_host_worker_failure_after_cleanup(monkeypatch):
+    engine, _transport, ring_engine = _engine_with_fake_ring()
+    deactivated = []
+
+    class _FailingHostEngine:
+        def __init__(self):
+            self.calls = []
+
+        def close_input(self):
+            self.calls.append("close_input")
+
+        def stop(self):
+            self.calls.append("stop")
+
+        def raise_if_failed(self):
+            self.calls.append("raise_if_failed")
+            raise RuntimeError("clickhouse worker failed")
+
+    host_engine = _FailingHostEngine()
+    engine._host_engine = host_engine
+    fake_transport_module = ModuleType("dmi.transport.ring")
+    fake_transport_module.deactivate = lambda: deactivated.append(True)
+    monkeypatch.setitem(sys.modules, "dmi.transport.ring", fake_transport_module)
+
+    with pytest.raises(RuntimeError, match="clickhouse worker failed"):
+        engine.close()
+
+    assert ring_engine.stop_calls == 1
+    assert deactivated == [True]
+    assert host_engine.calls == ["close_input", "stop", "raise_if_failed"]
+    assert engine._ring_engine is None
+    assert engine._ring_transport is None
+    assert engine._host_engine is None
+
+
+def test_close_preserves_ring_failure_while_cleaning_up_host(monkeypatch):
+    engine, _transport, ring_engine = _engine_with_fake_ring(fail_stop=True)
+    deactivated = []
+
+    class _HostEngine:
+        def __init__(self):
+            self.calls = []
+
+        def close_input(self):
+            self.calls.append("close_input")
+
+        def stop(self):
+            self.calls.append("stop")
+
+        def raise_if_failed(self):
+            self.calls.append("raise_if_failed")
+
+    host_engine = _HostEngine()
+    engine._host_engine = host_engine
+    fake_transport_module = ModuleType("dmi.transport.ring")
+    fake_transport_module.deactivate = lambda: deactivated.append(True)
+    monkeypatch.setitem(sys.modules, "dmi.transport.ring", fake_transport_module)
+
+    with pytest.raises(RuntimeError, match="ring drain failed"):
+        engine.close()
+
+    assert ring_engine.stop_calls == 1
+    assert deactivated == [True]
+    assert host_engine.calls == ["close_input", "stop", "raise_if_failed"]
+    assert engine._ring_engine is None
+    assert engine._ring_transport is None
+    assert engine._host_engine is None
+
+
 def test_replacing_disabled_ring_restores_native_null_mode(monkeypatch):
     engine, _transport, old_ring = _engine_with_fake_ring(null_offload=True)
     new_ring = _FakeRingEngine()

From ee1d71a59c391e67b5ff9e9e3676695afb4c68cb Mon Sep 17 00:00:00 2001
From: Alan Liu <zaoxingliu@gmail.com>
Date: Mon, 24 Aug 2026 22:38:14 -0400
Subject: [PATCH 2/5] Fix CPU test resource markers

---
 tests/test_cpu_native_build.py | 11 ++++++++---
 1 file changed, 8 insertions(+), 3 deletions(-)

diff --git a/tests/test_cpu_native_build.py b/tests/test_cpu_native_build.py
index f99abf60f..89bd1a0e1 100644
--- a/tests/test_cpu_native_build.py
+++ b/tests/test_cpu_native_build.py
@@ -9,9 +9,7 @@
 import pytest
 
 
-pytestmark = pytest.mark.cpu
-
-
+@pytest.mark.cpu
 def test_host_build_plan_has_no_cuda_toolchain_or_libraries():
     root = Path(__file__).resolve().parents[1]
     result = subprocess.run(
@@ -31,6 +29,7 @@ def test_host_build_plan_has_no_cuda_toolchain_or_libraries():
         assert forbidden not in output
 
 
+@pytest.mark.cpu
 def test_host_export_falls_back_to_cpu_backend(monkeypatch):
     from dmi.transport import native
 
@@ -50,6 +49,7 @@ def load_named(name):
     assert calls == ["_native_backend", "_host_backend"]
 
 
+@pytest.mark.native_backend
 def test_clickhouse_stage_has_bounded_batching_defaults():
     from dmi.transport.native import ClickHouseClientConfig, StageConfig
 
@@ -65,6 +65,7 @@ def test_clickhouse_stage_has_bounded_batching_defaults():
     assert queue.high_watermark_size == 512 * 1024**2
 
 
+@pytest.mark.native_backend
 def test_clickhouse_client_exposes_socket_timeouts_and_worker_metrics():
     from dmi.transport.native import ClickHouseClientConfig, DMXHostEngine, StageConfig
 
@@ -81,6 +82,7 @@ def test_clickhouse_client_exposes_socket_timeouts_and_worker_metrics():
     assert [worker.worker_index for worker in metrics.workers] == [0, 1, 2]
 
 
+@pytest.mark.native_backend
 def test_engine_metrics_follow_mutated_stage_parallelism():
     from dmi.transport.native import ClickHouseClientConfig, DMXHostEngine, StageConfig
 
@@ -93,6 +95,7 @@ def test_engine_metrics_follow_mutated_stage_parallelism():
     assert [worker.worker_index for worker in metrics.workers] == [0, 1, 2]
 
 
+@pytest.mark.cpu
 def test_ring_export_requires_full_backend(monkeypatch):
     from dmi.transport import native
 
@@ -110,6 +113,7 @@ def load_named(name):
     assert calls == ["_native_backend"]
 
 
+@pytest.mark.cpu
 def test_v1_host_export_does_not_load_ring_backend(monkeypatch):
     import dmi.api.v1 as api
     from dmi.transport import native
@@ -128,6 +132,7 @@ def test_v1_host_export_does_not_load_ring_backend(monkeypatch):
             api.__dict__["DMXHostEngine"] = cached
 
 
+@pytest.mark.cpu
 def test_v1_model_shape_contract_does_not_load_ring_backend():
     import dmi.api.v1 as api
 

From e99b296c25a6ff97a50d865f005947eee3b4549c Mon Sep 17 00:00:00 2001
From: Alan Liu <zaoxingliu@gmail.com>
Date: Mon, 24 Aug 2026 22:46:30 -0400
Subject: [PATCH 3/5] Limit PR to CPU ClickHouse work

---
 docs/clickhouse-offload-pipeline.html | 121 ++++++-------------
 docs/config.md                        |  19 ++-
 docs/integration-api-v1.md            |   2 +-
 native/csrc/bindings.cpp              |   9 +-
 native/csrc/ring/drain_thread.cpp     | 166 +++++++-------------------
 native/csrc/ring/drain_thread.h       |  29 +----
 native/csrc/ring/p2p_thread.cpp       |  76 ++++++++----
 native/csrc/ring/ring_config.h        |   5 +-
 native/csrc/ring/ring_engine.cu       |  39 +-----
 native/csrc/ring/ring_engine.h        |   2 -
 native/csrc/ring/ring_engine_py.cu    |  80 ++++++++-----
 native/csrc/ring/ring_engine_py.h     |  15 ++-
 native/csrc/ring/task_entry.h         |   2 +-
 native/csrc/ring/task_ring.cuh        |   6 +-
 src/dmi/engine.py                     |  42 ++-----
 src/dmi/hooks/point.py                |   8 +-
 src/dmi/transport/ring.py             |   4 +-
 tests/native/ring/Makefile            |   3 +-
 tests/native/ring/test_producer.cu    |  27 -----
 tests/native/ring/test_ring_engine.cu | 141 ++--------------------
 tests/test_engine_runtime_api.py      |  97 +--------------
 21 files changed, 259 insertions(+), 634 deletions(-)

diff --git a/docs/clickhouse-offload-pipeline.html b/docs/clickhouse-offload-pipeline.html
index 42c084cc9..2152136e2 100644
--- a/docs/clickhouse-offload-pipeline.html
+++ b/docs/clickhouse-offload-pipeline.html
@@ -3,7 +3,7 @@
 <head>
   <meta charset="utf-8">
   <meta name="viewport" content="width=device-width, initial-scale=1">
-  <title>DMI ClickHouse offload pipeline</title>
+  <title>DMI host-to-ClickHouse benchmark</title>
   <style>
     :root {
       color-scheme: light dark;
@@ -12,10 +12,8 @@
       --text: #eef7ff;
       --muted: #a7bdd2;
       --line: #35526c;
-      --gpu: #8aa7ff;
       --host: #55d6be;
       --db: #ffc857;
-      --error: #ff7b8b;
     }
     * { box-sizing: border-box; }
     body {
@@ -24,26 +22,21 @@
       color: var(--text);
       font: 16px/1.5 system-ui, sans-serif;
     }
-    main { width: min(1180px, 94vw); margin: auto; padding: 48px 0 64px; }
+    main { width: min(1120px, 94vw); margin: auto; padding: 48px 0 64px; }
     h1 { margin: 0 0 8px; font-size: clamp(2rem, 5vw, 3.4rem); }
     h2 { margin: 0 0 8px; font-size: 1.05rem; }
     p { color: var(--muted); }
-    code { color: var(--db); }
-    .flow {
-      display: grid;
-      grid-template-columns: repeat(5, 1fr);
-      gap: 14px;
-      margin: 36px 0 24px;
-    }
+    .flow, .measure { display: grid; gap: 14px; margin-top: 28px; }
+    .flow { grid-template-columns: repeat(4, 1fr); }
+    .measure { grid-template-columns: repeat(3, 1fr); }
     .card {
-      min-height: 190px;
+      min-height: 180px;
       padding: 20px;
       border: 1px solid var(--line);
       border-radius: 16px;
       background: color-mix(in srgb, var(--panel) 94%, transparent);
       box-shadow: 0 16px 40px #0004;
     }
-    .gpu { border-color: var(--gpu); }
     .host { border-color: var(--host); }
     .db { border-color: var(--db); }
     .tag {
@@ -55,97 +48,61 @@
       color: var(--muted);
       font-size: .8rem;
     }
-    .rules { display: grid; grid-template-columns: 1fr 1fr; gap: 14px; }
-    .measure { display: grid; grid-template-columns: repeat(3, 1fr); gap: 14px; margin-top: 14px; }
-    .error { border-color: var(--error); }
-    ul { margin: 10px 0 0; padding-left: 20px; color: var(--muted); }
     @media (max-width: 900px) {
-      .flow { grid-template-columns: 1fr; }
+      .flow, .measure { grid-template-columns: 1fr; }
       .card { min-height: 0; }
-      .rules { grid-template-columns: 1fr; }
-      .measure { grid-template-columns: 1fr; }
     }
   </style>
 </head>
 <body>
   <main>
-    <h1>GPU capture to ClickHouse</h1>
-    <p>The offload path overlaps GPU capture, host transfer, reconstruction,
-      and native ClickHouse insertion while keeping ownership and failures explicit.</p>
+    <h1>CPU host to ClickHouse</h1>
+    <p>The benchmark isolates ClickHouse ingestion from model execution and
+      accelerator transport. Deterministic CPU tensors exercise the same native
+      host queue and ClickHouse clients used by DMI.</p>
 
-    <section class="flow" aria-label="DMI offload stages">
-      <article class="card gpu">
-        <h2>1. GPU producer</h2>
-        <p>Copies an activation into the payload ring and publishes its descriptor
-          with a system-scope release fence.</p>
-        <span class="tag">device memory</span><span class="tag">CUDA stream</span>
-      </article>
-      <article class="card gpu">
-        <h2>2. Capacity owner</h2>
-        <p>Atomically checks payload, pinned-staging, and task-slot capacity before
-          reserving the next step or eager hook.</p>
-        <span class="tag">one snapshot</span><span class="tag">backpressure</span>
+    <section class="flow" aria-label="Host benchmark stages">
+      <article class="card host">
+        <h2>1. Synthetic producer</h2>
+        <p>Reuses a seeded pool of contiguous CPU tensors and submits a fixed
+          logical payload without model or device variance.</p>
+        <span class="tag">CPU tensors</span><span class="tag">repeatable</span>
       </article>
       <article class="card host">
-        <h2>3. Drain thread</h2>
-        <p>Batches checked D2H copies into pinned memory. Ring space is committed
-          only after the CUDA copy stream succeeds.</p>
-        <span class="tag">pinned ring</span><span class="tag">100 ms flush</span>
+        <h2>2. Bounded queue</h2>
+        <p>Applies byte and row watermarks, finite admission waits, and
+          configurable batch size and linger thresholds.</p>
+        <span class="tag">backpressure</span><span class="tag">16–64 MiB batches</span>
       </article>
       <article class="card host">
-        <h2>4. P2P thread</h2>
-        <p>Reconstructs pageable tensors, validates metadata, slices requests, and
-          submits rows to the host pipeline.</p>
-        <span class="tag">ATen CPU</span><span class="tag">request slices</span>
+        <h2>3. Native clients</h2>
+        <p>Each worker owns one ClickHouse connection. Readiness, per-worker
+          rows and bytes, insert time, and realized concurrency are recorded.</p>
+        <span class="tag">1 / 2 / 4 / 8 workers</span><span class="tag">socket timeouts</span>
       </article>
       <article class="card db">
-        <h2>5. ClickHouse stage</h2>
-        <p>Targets 16 MiB batches with a 50 ms linger and bounded queueing. Each
-          stage initializes its own destination once.</p>
-        <span class="tag">native protocol</span><span class="tag">512 MiB queue</span>
-      </article>
-    </section>
-
-    <section class="rules">
-      <article class="card">
-        <h2>Ownership rules</h2>
-        <ul>
-          <li>GPU payload is reusable only after successful D2H completion.</li>
-          <li>Pinned bytes are reusable only after pageable reconstruction.</li>
-          <li>ClickHouse column views retain tensor storage through insertion.</li>
-        </ul>
-      </article>
-      <article class="card error">
-        <h2>Failure rules</h2>
-        <ul>
-          <li>The first CUDA, reconstruction, submission, or worker error wins.</li>
-          <li>New reservations stop after failure.</li>
-          <li>Shutdown cleans every stage, then raises the preserved error.</li>
-        </ul>
+        <h2>4. ClickHouse</h2>
+        <p>Stores tensor metadata and binary payloads in a MergeTree table,
+          then verifies row and byte counts after the queue drains.</p>
+        <span class="tag">native protocol</span><span class="tag">verified output</span>
       </article>
     </section>
 
-    <h2 style="margin-top: 30px">CPU ingestion scaling</h2>
-    <p>Parallelism is useful only when requested workers become ready, perform
-      overlapping inserts, and move the ClickHouse throughput ceiling.</p>
-    <section class="measure" aria-label="CPU ingestion measurement stages">
+    <section class="measure" aria-label="Scaling evidence">
       <article class="card host">
-        <h2>Ready boundary</h2>
-        <p>The steady-state clock starts after every worker has connected,
-          selected the database, and applied its session settings.</p>
-        <span class="tag">startup separated</span><span class="tag">finite socket timeout</span>
+        <h2>Steady-state boundary</h2>
+        <p>Throughput starts only after every worker connects, selects the
+          database, and applies session settings.</p>
       </article>
       <article class="card host">
-        <h2>Realized concurrency</h2>
-        <p>Snapshots retain per-worker batches, rows, bytes, insert time, and
-          peak simultaneous native inserts.</p>
-        <span class="tag">worker distribution</span><span class="tag">client peak</span>
+        <h2>Repeated scaling sweep</h2>
+        <p>Shuffled trials retain raw results and summarize median throughput,
+          variance, speedup, and marginal gain.</p>
       </article>
       <article class="card db">
-        <h2>Saturation evidence</h2>
-        <p>Repeated shuffled worker sweeps combine throughput variance with
-          sampled active inserts, queries, merges, and connections.</p>
-        <span class="tag">1 → 2 → 4 → 8</span><span class="tag">raw trials retained</span>
+        <h2>Server evidence</h2>
+        <p>Active inserts, queries, merges, connections, CPU, I/O wait, memory,
+          and network counters are sampled over the measured interval.</p>
       </article>
     </section>
   </main>
diff --git a/docs/config.md b/docs/config.md
index 4d13678a1..7cc673a1c 100644
--- a/docs/config.md
+++ b/docs/config.md
@@ -38,22 +38,20 @@ ring. Force flush at 100% capacity is always active (prevents deadlock).
 | `drain_flush_payload_ratio` | `float` | 0.5 | Flush when scanned payload bytes >= this fraction of `payload_ring_bytes`. 0 = disabled. |
 | `drain_flush_entry_threshold` | `uint64_t` | 0 | Flush after N entries ready. 0 = disabled. |
 | `drain_flush_byte_threshold` | `uint64_t` | 0 | Flush after N payload bytes ready. 0 = disabled. |
-| `drain_flush_timeout_us` | `uint64_t` | 100000 | Flush completed tensors after this many microseconds. 0 disables the timer. |
+| `drain_flush_timeout_us` | `uint64_t` | 0 | If a complete tensor has been pending for longer than this many microseconds, flush unconditionally. 0 = disabled. |
 
-By default, the drain thread flushes after 100 ms or at 50% payload-ring usage.
-If all thresholds and the timeout are explicitly set to 0, it flushes only when
-the ring is full or at `stop()` time.
+By default, timeout-based flushing is disabled and the drain thread flushes at
+50% payload-ring usage. If `drain_flush_payload_ratio` and all other thresholds
+are explicitly set to 0, the drain thread only flushes when the ring is 100%
+full or at `stop()` time.
 
 ## P2P Thread / Output
 
 | Parameter | Type | Default | Description |
 |-----------|------|---------|-------------|
 | `clone_slices` | `bool` | false | Clone per-request slices before submitting to the host engine. When true (and batch > 1), each slice is an independent tensor so the full assembled tensor can be freed immediately. When false, slices are views that keep the full tensor alive until consumed. |
-| `insert_queue_max_bytes` | `uint64_t` | 4 GiB | Reserved; does not configure the host queue. |
-| `insert_queue_max_items` | `uint64_t` | 65536 | Reserved; does not configure the host queue. |
-
-Configure ClickHouse batching and backpressure through
-`StageConfig.input_queue`.
+| `insert_queue_max_bytes` | `uint64_t` | 4 GiB | ClickHouse insert queue byte limit. The p2p thread blocks when the queue is full. |
+| `insert_queue_max_items` | `uint64_t` | 65536 | ClickHouse insert queue item-count limit. |
 
 ## Constants (not configurable)
 
@@ -61,7 +59,7 @@ Configure ClickHouse batching and backpressure through
 |----------|-------|----------|-------------|
 | `PAYLOAD_ALIGN` | 16 bytes | `ring_config.h` | Payload allocation alignment. Every reservation is rounded up to this for vectorized uint4 D2D copies. `payload_ring_bytes` must be a multiple of this. |
 | `READY_SEQ_SENTINEL` | `UINT64_MAX` | `task_entry.h` | Sentinel value for `TaskEntry::ready_seq` (slot not yet published). |
-| `TaskEntry` size | 64 bytes | `task_entry.h` | Fixed slot size, `alignas(64)` for cache-line isolation. |
+| `TaskEntry` size | 128 bytes | `task_entry.h` | Fixed slot size, `alignas(128)` for cache-line isolation. |
 
 ## Python Usage
 
@@ -76,6 +74,7 @@ cfg.drain_poll_timeout_us = 100
 cfg.drain_flush_entry_threshold = 64
 cfg.drain_flush_timeout_us = 1000
 cfg.clone_slices = False
+cfg.insert_queue_max_items = 4096
 
 engine = RingEngine(cfg, host_engine)
 engine.init(stream_handle)
diff --git a/docs/integration-api-v1.md b/docs/integration-api-v1.md
index 77310d9b8..deb25fd33 100644
--- a/docs/integration-api-v1.md
+++ b/docs/integration-api-v1.md
@@ -814,7 +814,7 @@ do not reconfigure it.
 | `drain_flush_payload_ratio` | `0.5` | Payload-capacity flush fraction. |
 | `drain_flush_entry_threshold` | `0` | Absolute ready-entry trigger; zero disables it. |
 | `drain_flush_byte_threshold` | `0` | Absolute ready-byte trigger; zero disables it. |
-| `drain_flush_timeout_us` | `100000` | Pending-data age trigger; zero disables it. |
+| `drain_flush_timeout_us` | `0` | Pending-data age trigger; zero disables it. |
 | `clone_slices` | `False` | Clone multi-request slices so full assembled tensors can be released sooner. |
 | `insert_queue_max_bytes` | `4 GiB` | Reserved field; current v1 does not apply it to host queue limits. |
 | `insert_queue_max_items` | `65536` | Reserved field; current v1 does not apply it to host queue limits. |
diff --git a/native/csrc/bindings.cpp b/native/csrc/bindings.cpp
index 8190cafae..7e33e5ed2 100644
--- a/native/csrc/bindings.cpp
+++ b/native/csrc/bindings.cpp
@@ -351,12 +351,13 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
       .def("staging_cap", &ring_py::RingEnginePy::staging_cap)
       .def("task_cap",    &ring_py::RingEnginePy::task_cap)
       .def("payload_tensor", &ring_py::RingEnginePy::payload_tensor)
-      // Safety-net surface (eager only). CPU-only and fast.
+      // Safety-net surface (eager only).  available_capacity() and
+      // reserve_one() are CPU-only and fast -- no GIL release needed.
       // flush_and_wait() blocks on cudaStreamSynchronize + drain flush --
       // GIL released so other Python threads aren't blocked.
-      .def("effective_capacity", &ring_py::RingEnginePy::effective_capacity)
-      .def("try_reserve_one",
-           &ring_py::RingEnginePy::try_reserve_one,
+      .def("available_capacity", &ring_py::RingEnginePy::available_capacity)
+      .def("reserve_one",
+           &ring_py::RingEnginePy::reserve_one,
            py::arg("nbytes"))
       .def("flush_and_wait",
            &ring_py::RingEnginePy::flush_and_wait,
diff --git a/native/csrc/ring/drain_thread.cpp b/native/csrc/ring/drain_thread.cpp
index a6f397a08..6fbcfa428 100644
--- a/native/csrc/ring/drain_thread.cpp
+++ b/native/csrc/ring/drain_thread.cpp
@@ -34,23 +34,16 @@ DrainThread::~DrainThread() noexcept {
 
 void DrainThread::start() {
     running_.store(true, std::memory_order_relaxed);
-    thread_ = std::thread([this] {
-        try {
-            loop();
-        } catch (...) {
-            report_failure(std::current_exception());
-        }
-    });
+    thread_ = std::thread([this] { loop(); });
 }
 
 void DrainThread::stop() {
-    running_.store(false, std::memory_order_relaxed);
+    if (!running_.exchange(false)) return;
     cv_.notify_all();
     if (thread_.joinable()) thread_.join();
 }
 
 void DrainThread::notify() {
-    rethrow_if_failed();
     {
         std::lock_guard<std::mutex> lk(mu_);
         notified_ = true;
@@ -65,7 +58,6 @@ void DrainThread::notify() {
 // cudaStreamSynchronize(main_stream) first so all GPU writes are visible.
 // ---------------------------------------------------------------------------
 void DrainThread::force_flush_and_wait() {
-    rethrow_if_failed();
     {
         std::lock_guard<std::mutex> lk(mu_);
         flush_requested_ = true;
@@ -76,37 +68,7 @@ void DrainThread::force_flush_and_wait() {
 
     // Block until drain thread completes the flush
     std::unique_lock<std::mutex> lk(mu_);
-    flush_done_cv_.wait(lk, [this] { return flush_done_ || has_failed(); });
-    lk.unlock();
-    rethrow_if_failed();
-}
-
-void DrainThread::report_failure(std::exception_ptr failure) noexcept {
-    {
-        std::lock_guard<std::mutex> lk(failure_mu_);
-        if (!failure_) failure_ = std::move(failure);
-    }
-    failed_.store(true, std::memory_order_release);
-    running_.store(false, std::memory_order_relaxed);
-    {
-        std::lock_guard<std::mutex> lk(mu_);
-        flush_done_ = true;
-    }
-    {
-        std::lock_guard<std::mutex> lk(pop_mu_);
-        p2p_stop_requested_ = true;
-    }
-    cv_.notify_all();
-    flush_done_cv_.notify_all();
-    pop_cv_.notify_all();
-    staging_cv_.notify_all();
-}
-
-void DrainThread::rethrow_if_failed() const {
-    if (!has_failed()) return;
-    std::lock_guard<std::mutex> lk(failure_mu_);
-    if (failure_) std::rethrow_exception(failure_);
-    throw std::runtime_error("DrainThread failed");
+    flush_done_cv_.wait(lk, [this] { return flush_done_; });
 }
 
 // ---------------------------------------------------------------------------
@@ -149,37 +111,22 @@ void DrainThread::notify_staging_freed_bytes(uint64_t nbytes) {
 }
 
 // ---------------------------------------------------------------------------
-// Capacity query
+// Capacity query accessors
 // ---------------------------------------------------------------------------
-CapacitySnapshot DrainThread::capacity_snapshot() {
-    rethrow_if_failed();
-    std::lock_guard<std::mutex> lk(mgmt_mu_);
-    return {
-        cpu_payload_head_,
-        cpu_payload_tail_committed_,
-        cpu_task_head_,
-        cpu_task_tail_,
-    };
+uint64_t DrainThread::cpu_payload_head() const {
+    return cpu_payload_head_;
 }
 
-bool DrainThread::try_reserve(uint64_t payload_bytes, uint32_t num_tasks,
-                              uint64_t payload_capacity,
-                              uint64_t task_capacity) {
-    rethrow_if_failed();
-    std::lock_guard<std::mutex> lk(mgmt_mu_);
-    const uint64_t payload_used =
-        cpu_payload_head_ - cpu_payload_tail_committed_;
-    const uint64_t tasks_used = cpu_task_head_ - cpu_task_tail_;
-    if (payload_used > payload_capacity || tasks_used > task_capacity) {
-        throw std::logic_error("DrainThread: capacity invariant violated");
-    }
-    if (payload_bytes > payload_capacity - payload_used ||
-        num_tasks > task_capacity - tasks_used) {
-        return false;
-    }
-    cpu_payload_head_ += payload_bytes;
-    cpu_task_head_ += num_tasks;
-    return true;
+uint64_t DrainThread::cpu_payload_tail_committed() const {
+    return cpu_payload_tail_committed_;
+}
+
+uint64_t DrainThread::cpu_task_head() const {
+    return cpu_task_head_;
+}
+
+uint64_t DrainThread::cpu_task_tail_committed() const {
+    return cpu_task_tail_;
 }
 
 // ---------------------------------------------------------------------------
@@ -187,7 +134,6 @@ bool DrainThread::try_reserve(uint64_t payload_bytes, uint32_t num_tasks,
 // Called from prepare_step after confirming space is available.
 // ---------------------------------------------------------------------------
 void DrainThread::reserve(uint64_t payload_bytes, uint32_t num_tasks) {
-    rethrow_if_failed();
     std::lock_guard<std::mutex> lk(mgmt_mu_);
     cpu_payload_head_ += payload_bytes;
     cpu_task_head_    += num_tasks;
@@ -197,7 +143,6 @@ void DrainThread::reserve(uint64_t payload_bytes, uint32_t num_tasks) {
 // submit_cpu_direct -- submit a CPU-direct tensor to drain -> p2p pipeline.
 // ---------------------------------------------------------------------------
 void DrainThread::submit_cpu_direct(at::Tensor cpu_tensor, uint64_t tensor_bytes) {
-    rethrow_if_failed();
     DrainTask task{};
     task.tensor_total_bytes = tensor_bytes;
     task.cpu_paged_tensor   = std::move(cpu_tensor);
@@ -218,7 +163,7 @@ void DrainThread::submit_cpu_direct(at::Tensor cpu_tensor, uint64_t tensor_bytes
 // ---------------------------------------------------------------------------
 void DrainThread::do_full_flush() {
     for (;;) {
-        uint64_t flush_count = 0, flush_bytes = 0, src_start = 0;
+        uint64_t flush_count = 0, flush_bytes = 0;
         {
             std::lock_guard<std::mutex> lk(mgmt_mu_);
             scan_ready();
@@ -230,20 +175,17 @@ void DrainThread::do_full_flush() {
                 flush_count++;
             }
             if (flush_count == 0) break;
-            src_start = cpu_payload_tail_;
+            flush_state_update(flush_count, flush_bytes);
         }
         {
             std::unique_lock<std::mutex> lk(staging_mu_);
-            staging_cv_.wait(lk, [&] {
-                return staging_.free_bytes() >= flush_bytes || has_failed();
-            });
+            staging_cv_.wait(lk, [&] { return staging_.free_bytes() >= flush_bytes; });
         }
-        rethrow_if_failed();
-        enqueue_d2h(flush_bytes, src_start);
+        enqueue_d2h(flush_bytes);
         sync_stream();
         {
             std::lock_guard<std::mutex> lk(mgmt_mu_);
-            flush_state_update(flush_count, flush_bytes);
+            cpu_payload_tail_committed_ = cpu_payload_tail_;
         }
         submit_to_p2p(flush_count, flush_bytes);
         {
@@ -285,7 +227,7 @@ void DrainThread::loop() {
             continue;  // skip normal sleep, re-check immediately
         }
 
-        uint64_t flush_count = 0, flush_bytes = 0, src_start = 0;
+        uint64_t flush_count = 0, flush_bytes = 0;
         bool needs_flush = false;
 
         {
@@ -306,7 +248,7 @@ void DrainThread::loop() {
                             (unsigned long)flush_bytes, (unsigned long)pending_entries_,
                             (unsigned long)staging_.free_bytes());
 
-                    src_start = cpu_payload_tail_;
+                    flush_state_update(flush_count, flush_bytes);
                     needs_flush = true;
                 }
             }
@@ -316,17 +258,16 @@ void DrainThread::loop() {
             {
                 std::unique_lock<std::mutex> lk(staging_mu_);
                 staging_cv_.wait(lk, [&] {
-                    return staging_.free_bytes() >= flush_bytes || has_failed();
+                    return staging_.free_bytes() >= flush_bytes;
                 });
             }
 
-            rethrow_if_failed();
-            enqueue_d2h(flush_bytes, src_start);
+            enqueue_d2h(flush_bytes);
             sync_stream();
 
             {
                 std::lock_guard<std::mutex> lk(mgmt_mu_);
-                flush_state_update(flush_count, flush_bytes);
+                cpu_payload_tail_committed_ = cpu_payload_tail_;
             }
 
             submit_to_p2p(flush_count, flush_bytes);
@@ -348,12 +289,7 @@ void DrainThread::loop() {
     }
 
     // Final flush
-    cudaError_t error = cudaDeviceSynchronize();
-    if (error != cudaSuccess) {
-        throw std::runtime_error(
-            std::string("DrainThread: cudaDeviceSynchronize failed: ") +
-            cudaGetErrorString(error));
-    }
+    cudaDeviceSynchronize();
     do_full_flush();
 }
 
@@ -418,23 +354,18 @@ void DrainThread::flush_state_update(uint64_t flush_count, uint64_t flush_bytes)
         ++cpu_task_tail_;
     }
     cpu_payload_tail_ += flush_bytes;
-    cpu_payload_tail_committed_ = cpu_payload_tail_;
 }
 
 void DrainThread::sync_stream() {
-    cudaError_t error = cudaStreamSynchronize(stream_);
-    if (error != cudaSuccess) {
-        throw std::runtime_error(
-            std::string("DrainThread: cudaStreamSynchronize failed: ") +
-            cudaGetErrorString(error));
-    }
+    cudaStreamSynchronize(stream_);
 }
 
 // ---------------------------------------------------------------------------
-void DrainThread::enqueue_d2h(uint64_t flush_bytes, uint64_t src_start) {
+void DrainThread::enqueue_d2h(uint64_t flush_bytes) {
     if (flush_bytes == 0) return;
     const uint64_t gpu_cap = ring_.payload_cap;
     const uint64_t stg_cap = staging_.capacity();
+    uint64_t src_start = cpu_payload_tail_ - flush_bytes;
     uint64_t gpu_cursor = src_start % gpu_cap;
     uint64_t stg_cursor = staging_.head() % stg_cap;
     uint64_t remaining  = flush_bytes;
@@ -453,9 +384,8 @@ void DrainThread::enqueue_d2h(uint64_t flush_bytes, uint64_t src_start) {
                         ring_.payload_buf + gpu_cursor,
                         chunk, cudaMemcpyDeviceToHost, stream_);
         if (err != cudaSuccess) {
-            throw std::runtime_error(
-                std::string("DrainThread: cudaMemcpyAsync failed: ") +
-                cudaGetErrorString(err));
+            RING_DBG("[enqueue_d2h] cudaMemcpyAsync FAILED: %s\n",
+                    cudaGetErrorString(err));
         }
         RING_DBG("[enqueue_d2h] chunk=%d enqueued OK\n", chunk_idx);
 
@@ -472,13 +402,7 @@ void DrainThread::enqueue_d2h(uint64_t flush_bytes, uint64_t src_start) {
 // ---------------------------------------------------------------------------
 void DrainThread::submit_to_p2p(uint64_t flush_count, uint64_t flush_bytes) {
     uint64_t cumulative = 0;
-    uint64_t staging_batch_start = 0;
-    {
-        std::lock_guard<std::mutex> lk(staging_mu_);
-        staging_batch_start = staging_.head();
-    }
-    std::vector<DrainTask> ready;
-    ready.reserve(flush_count);
+    const uint64_t staging_batch_start = staging_.head();
 
     for (uint64_t i = 0; i < flush_count; ++i) {
         const TaskEntry& ec = scanned_[i];
@@ -506,22 +430,18 @@ void DrainThread::submit_to_p2p(uint64_t flush_count, uint64_t flush_bytes) {
             cumulative += alloc;
         }
 
-        ready.push_back(std::move(task));
+        {
+            std::lock_guard<std::mutex> lk(queue_mu_);
+            task_queue_.push_back(std::move(task));
+        }
+        {
+            std::lock_guard<std::mutex> lk(pop_mu_);
+            can_pop_count_ += 1;
+        }
+        pop_cv_.notify_one();
     }
 
-    {
-        std::lock_guard<std::mutex> lk(queue_mu_);
-        for (auto& task : ready) task_queue_.push_back(std::move(task));
-    }
-    {
-        std::lock_guard<std::mutex> lk(staging_mu_);
-        staging_.advance_head(flush_bytes);
-    }
-    {
-        std::lock_guard<std::mutex> lk(pop_mu_);
-        can_pop_count_ += flush_count;
-    }
-    pop_cv_.notify_one();
+    staging_.advance_head(flush_bytes);
 }
 
 // ---------------------------------------------------------------------------
diff --git a/native/csrc/ring/drain_thread.h b/native/csrc/ring/drain_thread.h
index d0c1071fa..66d03780b 100644
--- a/native/csrc/ring/drain_thread.h
+++ b/native/csrc/ring/drain_thread.h
@@ -22,20 +22,12 @@
 #include <atomic>
 #include <condition_variable>
 #include <deque>
-#include <exception>
 #include <mutex>
 #include <thread>
 #include <vector>
 
 namespace ring {
 
-struct CapacitySnapshot {
-    uint64_t payload_head;
-    uint64_t payload_tail_committed;
-    uint64_t task_head;
-    uint64_t task_tail_committed;
-};
-
 class DrainThread {
 public:
     DrainThread(RingState& rs, PinnedStaging& staging, const RingConfig& cfg);
@@ -65,18 +57,13 @@ class DrainThread {
 
     void notify_staging_freed_bytes(uint64_t nbytes);
 
-    void report_failure(std::exception_ptr failure) noexcept;
-    void rethrow_if_failed() const;
-    bool has_failed() const noexcept {
-        return failed_.load(std::memory_order_acquire);
-    }
-
     bool is_running() const { return running_.load(std::memory_order_relaxed); }
 
-    CapacitySnapshot capacity_snapshot();
-
-    bool try_reserve(uint64_t payload_bytes, uint32_t num_tasks,
-                     uint64_t payload_capacity, uint64_t task_capacity);
+    // Capacity query accessors (called from RingEnginePy::prepare_step).
+    uint64_t cpu_payload_head() const;
+    uint64_t cpu_payload_tail_committed() const;
+    uint64_t cpu_task_head() const;
+    uint64_t cpu_task_tail_committed() const;
 
     // Pre-allocate ring space for the next step's producer kernels.
     // Advances cpu_payload_head_ and cpu_task_head_ under mgmt_mu_.
@@ -128,10 +115,6 @@ class DrainThread {
     std::mutex              staging_mu_;
     std::condition_variable staging_cv_;
 
-    mutable std::mutex      failure_mu_;
-    std::exception_ptr      failure_;
-    std::atomic<bool>       failed_{false};
-
     void loop();
 
     // Drain all pending entries -- called by the drain thread when
@@ -144,7 +127,7 @@ class DrainThread {
     void flush_state_update(uint64_t flush_count, uint64_t flush_bytes);
 
     void sync_stream();
-    void enqueue_d2h(uint64_t flush_bytes, uint64_t src_start);
+    void enqueue_d2h(uint64_t flush_bytes);
 
     // Split into two: submit_to_p2p pushes DrainTasks to the p2p queue
     // (uses queue_mu_/pop_mu_, NOT mgmt_mu_).  trim_scanned updates
diff --git a/native/csrc/ring/p2p_thread.cpp b/native/csrc/ring/p2p_thread.cpp
index 60e4beefe..f68426297 100644
--- a/native/csrc/ring/p2p_thread.cpp
+++ b/native/csrc/ring/p2p_thread.cpp
@@ -6,9 +6,10 @@
 #include "pinned_staging.h"
 
 #include <ATen/ATen.h>
+#include <cstdio>
 #include <cstring>
 #include <exception>
-#include <stdexcept>
+#include <mutex>
 
 namespace ring {
 
@@ -16,6 +17,36 @@ namespace ring {
 // ATen helpers (no GIL required for CPU tensors)
 // ---------------------------------------------------------------------------
 
+static std::once_flag g_submit_failure_log_once;
+
+static void log_submit_failure_once(
+    const std::string& model_id,
+    const std::string& req_id,
+    const std::string& act_name,
+    int32_t layer_no,
+    int32_t shard_rank,
+    int32_t start_token,
+    int32_t end_token,
+    const char* error)
+{
+    std::call_once(g_submit_failure_log_once, [&] {
+        fprintf(stderr,
+                "[DMI][P2P] WARN: failed to submit tensor slice to host "
+                "engine; suppressing further submit errors. model_id=%s "
+                "request_id=%s act_name=%s layer_no=%d shard_rank=%d "
+                "token_range=[%d,%d) error=\"%s\"\n",
+                model_id.c_str(),
+                req_id.c_str(),
+                act_name.c_str(),
+                layer_no,
+                shard_rank,
+                start_token,
+                end_token,
+                error ? error : "unknown");
+        fflush(stderr);
+    });
+}
+
 // Build ClickHouse act_name from hook_type.
 // Per-layer: "blocks.<hook_type_name>"  (e.g. "blocks.attn.hook_pattern")
 // Global:    "<hook_type_name>"         (e.g. "hook_embed", "token_ids")
@@ -101,13 +132,7 @@ P2PThread::~P2PThread() noexcept {
 }
 
 void P2PThread::start() {
-    thread_ = std::thread([this] {
-        try {
-            loop();
-        } catch (...) {
-            drain_.report_failure(std::current_exception());
-        }
-    });
+    thread_ = std::thread([this] { loop(); });
 }
 
 void P2PThread::stop() {
@@ -165,16 +190,14 @@ void P2PThread::process(std::vector<DrainTask>& tasks) {
 void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_task) {
     ring_py::TensorMeta meta;
     if (!fifo_.pop(meta)) {
-        throw std::runtime_error("P2PThread: tensor metadata queue is empty");
+        return;
     }
 
     // Get step context -- pop from context queue if this is the first
     // hook in a new step (current_ctx_ is null).
     if (!current_ctx_) {
         current_ctx_ = fifo_.pop_context();
-        if (!current_ctx_) {
-            throw std::runtime_error("P2PThread: step context queue is empty");
-        }
+        if (!current_ctx_) return;  // no context available
     }
 
     if (meta.shape.empty() || first_task.tensor_total_bytes == 0) {
@@ -214,11 +237,11 @@ void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_ta
     }
 
     if (static_cast<uint64_t>(expected_bytes) != first_task.tensor_total_bytes) {
-        throw std::runtime_error(
-            "P2PThread: shape/bytes mismatch for " +
-            std::string(ring_py::hook_type_name(meta.hook_type)) +
-            ": expected=" + std::to_string(expected_bytes) +
-            " actual=" + std::to_string(first_task.tensor_total_bytes));
+        fprintf(stderr, "[p2p] WARN: shape/bytes mismatch: expected=%ld actual=%lu hook=%s\n",
+                (long)expected_bytes, (unsigned long)first_task.tensor_total_bytes,
+                ring_py::hook_type_name(meta.hook_type));
+        if (meta.last_in_step) { delete current_ctx_; current_ctx_ = nullptr; }
+        return;
     }
 
     tensor = tensor.view(dtype).reshape(
@@ -283,10 +306,21 @@ void P2PThread::do_post_processing(at::Tensor& tensor, const DrainTask& first_ta
             slice = slice.clone();
         }
 
-        submit_fn_(current_ctx_->model_id, shard_rank,
-                   req.req_id, act_name, meta.layer_no,
-                   db_start, db_end,
-                   std::move(slice));
+        try {
+            submit_fn_(current_ctx_->model_id, shard_rank,
+                       req.req_id, act_name, meta.layer_no,
+                       db_start, db_end,
+                       std::move(slice));
+        } catch (const std::exception& e) {
+            log_submit_failure_once(current_ctx_->model_id, req.req_id,
+                                    act_name, meta.layer_no, shard_rank,
+                                    db_start, db_end, e.what());
+        } catch (...) {
+            log_submit_failure_once(current_ctx_->model_id, req.req_id,
+                                    act_name, meta.layer_no, shard_rank,
+                                    db_start, db_end,
+                                    "unknown non-std exception");
+        }
     }
 
     // Last hook in step -- free context
diff --git a/native/csrc/ring/ring_config.h b/native/csrc/ring/ring_config.h
index 9e6d6d805..6f02fc875 100644
--- a/native/csrc/ring/ring_config.h
+++ b/native/csrc/ring/ring_config.h
@@ -36,7 +36,7 @@ struct DrainFlushConfig {
 
     // Time-based flush: if a complete tensor has been pending for longer
     // than this many microseconds, flush unconditionally.  0 = disabled.
-    uint64_t timeout_us = 100000;
+    uint64_t timeout_us = 0;
 };
 
 // ---------------------------------------------------------------------------
@@ -65,7 +65,8 @@ struct RingConfig {
     // that keep the full tensor alive until consumed.
     bool clone_slices = false;
 
-    // Reserved for compatibility. StageConfig.input_queue owns host limits.
+    // ClickHouse insert queue limits (host engine).
+    // P2p thread blocks on submit_direct() when queue is full.
     uint64_t insert_queue_max_bytes = 4096ULL * 1024 * 1024;  // 4 GiB
     uint64_t insert_queue_max_items = 65536;
 
diff --git a/native/csrc/ring/ring_engine.cu b/native/csrc/ring/ring_engine.cu
index 3bd876671..2b26cf903 100644
--- a/native/csrc/ring/ring_engine.cu
+++ b/native/csrc/ring/ring_engine.cu
@@ -3,9 +3,7 @@
 
 #include "ring_engine.h"
 
-#include <exception>
 #include <stdexcept>
-#include <string>
 
 namespace ring {
 
@@ -50,49 +48,20 @@ void RingEngine::init(cudaStream_t stream) {
 }
 
 void RingEngine::start() {
-    if (started_.exchange(true)) return;
     drain_->start();
     p2p_->start();
 }
 
 void RingEngine::stop() {
-    if (!started_.exchange(false)) {
-        drain_->rethrow_if_failed();
-        return;
-    }
+    // Guard against double-stop (benchmark _timed_close + engine.close).
+    if (!drain_->is_running()) return;
 
-    std::exception_ptr failure;
-    auto capture = [&failure] {
-        if (!failure) failure = std::current_exception();
-    };
+    cudaDeviceSynchronize();
+    drain_->force_flush_and_wait();
 
-    if (!drain_->has_failed()) {
-        cudaError_t error = cudaDeviceSynchronize();
-        if (error != cudaSuccess) {
-            try {
-                throw std::runtime_error(
-                    std::string("RingEngine: cudaDeviceSynchronize failed: ") +
-                    cudaGetErrorString(error));
-            } catch (...) {
-                capture();
-            }
-        } else {
-            try {
-                drain_->force_flush_and_wait();
-            } catch (...) {
-                capture();
-            }
-        }
-    }
     drain_->stop();
     drain_->signal_p2p_stop();
     p2p_->stop();
-    try {
-        drain_->rethrow_if_failed();
-    } catch (...) {
-        capture();
-    }
-    if (failure) std::rethrow_exception(failure);
 }
 
 }  // namespace ring
diff --git a/native/csrc/ring/ring_engine.h b/native/csrc/ring/ring_engine.h
index 64ce757d4..4a56a5f9a 100644
--- a/native/csrc/ring/ring_engine.h
+++ b/native/csrc/ring/ring_engine.h
@@ -7,7 +7,6 @@
 #include "p2p_thread.h"
 #include "tensor_meta.h"
 
-#include <atomic>
 #include <memory>
 #include <vector>
 
@@ -39,7 +38,6 @@ class RingEngine {
     PinnedStaging   staging_;
     std::unique_ptr<DrainThread>  drain_;
     std::unique_ptr<P2PThread>    p2p_;
-    std::atomic<bool>             started_{false};
 };
 
 }  // namespace ring
diff --git a/native/csrc/ring/ring_engine_py.cu b/native/csrc/ring/ring_engine_py.cu
index 40998ba13..c91c6d830 100644
--- a/native/csrc/ring/ring_engine_py.cu
+++ b/native/csrc/ring/ring_engine_py.cu
@@ -11,8 +11,6 @@
 #include "ring/producer.cuh"
 #include "ring/ring_debug.h"
 #include <ATen/cuda/CUDAContext.h>  // at::cuda::getCurrentCUDAStream
-#include <stdexcept>
-#include <string>
 
 // Forward-declare symbols from producer.cu
 namespace ring {
@@ -21,14 +19,6 @@ void set_ring_null_mode(bool enabled);
 
 namespace ring_py {
 
-static void check_cuda(cudaError_t error, const char* operation) {
-    if (error != cudaSuccess) {
-        throw std::runtime_error(
-            std::string("RingEngine: ") + operation + " failed: " +
-            cudaGetErrorString(error));
-    }
-}
-
 // ---------------------------------------------------------------------------
 struct RingEnginePy::Impl {
     TensorMetaFifo   fifo;
@@ -53,7 +43,7 @@ struct RingEnginePy::Impl {
     {
         const auto& state = engine.ring_state();
         int dev_idx = 0;
-        check_cuda(cudaGetDevice(&dev_idx), "cudaGetDevice");
+        cudaGetDevice(&dev_idx);
         payload_view = at::from_blob(
             state.payload_buf,
             {static_cast<int64_t>(state.payload_cap)},
@@ -107,9 +97,9 @@ void RingEnginePy::set_null_mode(bool enabled) {
     // NOT synchronize with PyTorch's non-blocking compute streams.  Sync
     // before to drain pending producer kernels that need the old value,
     // and after to ensure the new value is visible before the next launch.
-    check_cuda(cudaDeviceSynchronize(), "cudaDeviceSynchronize");
+    cudaDeviceSynchronize();
     ring::set_ring_null_mode(enabled);
-    check_cuda(cudaDeviceSynchronize(), "cudaDeviceSynchronize");
+    cudaDeviceSynchronize();
 }
 
 
@@ -184,8 +174,8 @@ void RingEnginePy::notify_drain() {
 // ---------------------------------------------------------------------------
 // prepare_step -- single Python->C++ call for pre-forward capacity check.
 //
-// Fast path (STEP_RING_OK): reads one locked capacity snapshot.
-// No CUDA stream resolution, sync, or flush.
+// Fast path (STEP_RING_OK): reads two uint64_t counters, returns immediately.
+// No stream resolution, no sync, no flush.
 //
 // Slow path (STEP_RING_FLUSHED / STEP_OVERSIZED): resolves the current CUDA
 // stream via at::cuda::getCurrentCUDAStream(), synchronises it, then asks the
@@ -234,23 +224,28 @@ int RingEnginePy::prepare_step(uint64_t step_total_bytes,
     // net starts firing.
     if (step_total_bytes > effective_cap || num_hooks > tcap) {
         cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-        check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
+        cudaStreamSynchronize(ms);
         drain.force_flush_and_wait();
         return STEP_OVERSIZED;
     }
 
-    if (drain.try_reserve(step_total_bytes, num_hooks, pcap, tcap)) {
-        return STEP_RING_OK;  // fast path -- no CUDA interaction
+    // Case A: step fits.  Check available space for BOTH payload AND tasks.
+    const uint64_t payload_avail = pcap -
+        (drain.cpu_payload_head() - drain.cpu_payload_tail_committed());
+    const uint64_t task_avail = tcap -
+        (drain.cpu_task_head() - drain.cpu_task_tail_committed());
+
+    if (step_total_bytes <= payload_avail && num_hooks <= task_avail) {
+        drain.reserve(step_total_bytes, num_hooks);
+        return STEP_RING_OK;  // fast path -- no CUDA or thread interaction
     }
 
     // Either payload or task ring full from prior steps.  Sync main
     // stream so all producer kernels finish writing, then flush.
     cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-    check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
+    cudaStreamSynchronize(ms);
     drain.force_flush_and_wait();
-    if (!drain.try_reserve(step_total_bytes, num_hooks, pcap, tcap)) {
-        throw std::logic_error("RingEngine: empty ring cannot fit step");
-    }
+    drain.reserve(step_total_bytes, num_hooks);
     return STEP_RING_FLUSHED;
 }
 
@@ -282,16 +277,43 @@ at::Tensor RingEnginePy::payload_tensor() const {
 // HookPoint.forward.  All three are called only when force_eager is active
 // (eager mode); never run during CUDA-graph capture or replay.
 //
-// try_reserve_one() checks payload, staging, and task capacity atomically.
+// Thread safety of the check-and-reserve pattern used by the safety net:
+//
+//   if nbytes <= available_capacity():
+//       reserve_one(nbytes)
+//
+// The main thread (this thread) is the only writer of cpu_payload_head_
+// (it advances only through reserve / reserve_one calls).  The drain
+// thread only ever advances cpu_payload_tail_committed_ forward as it
+// frees ring space.  Between the check and the reserve:
+//   - tail may move forward (drain freed more): actual available at
+//     reserve time is >= what we observed.
+//   - head is unchanged (single-threaded writer).
+// So the check's "fits" decision remains valid at reserve time.  No extra
+// locking around the pair is required.
+//
+// Within available_capacity(), the two accessor calls happen under
+// separate mutex acquires (drain.cpu_payload_head() and
+// drain.cpu_payload_tail_committed() each take mgmt_mu_ internally).
+// The observed snapshot is non-atomic: if drain advances tail between
+// the two reads, available_observed = pcap - head + tail_later, which
+// is >= the true available at the time of the head read.  That is, the
+// non-atomicity errs on the "over-estimate available" side -- the
+// reserve will still succeed because the actual ring state has at least
+// as much room as we computed.
 // ---------------------------------------------------------------------------
 
-uint64_t RingEnginePy::effective_capacity() const {
-    return std::min(impl_->engine.payload_cap(), impl_->engine.staging_cap());
+uint64_t RingEnginePy::available_capacity() const {
+    auto& drain = impl_->engine.drain_thread();
+    const uint64_t pcap = impl_->engine.payload_cap();
+    return pcap - (drain.cpu_payload_head() - drain.cpu_payload_tail_committed());
 }
 
-bool RingEnginePy::try_reserve_one(uint64_t nbytes) {
-    return impl_->engine.drain_thread().try_reserve(
-        nbytes, 1, effective_capacity(), impl_->engine.task_cap());
+// Per-hook reservation: claim nbytes of payload + 1 task entry for an
+// upcoming producer kernel launch.  Caller must have checked
+// available_capacity() first.  drain.reserve takes mgmt_mu_ internally.
+void RingEnginePy::reserve_one(uint64_t nbytes) {
+    impl_->engine.drain_thread().reserve(nbytes, 1);
 }
 
 // Synchronise the current CUDA stream so all queued producer kernels
@@ -300,7 +322,7 @@ bool RingEnginePy::try_reserve_one(uint64_t nbytes) {
 // Python binding releases the GIL.
 void RingEnginePy::flush_and_wait() {
     cudaStream_t ms = at::cuda::getCurrentCUDAStream().stream();
-    check_cuda(cudaStreamSynchronize(ms), "cudaStreamSynchronize");
+    cudaStreamSynchronize(ms);
     impl_->engine.drain_thread().force_flush_and_wait();
 }
 
diff --git a/native/csrc/ring/ring_engine_py.h b/native/csrc/ring/ring_engine_py.h
index 9f10f1ceb..e96e1e4a6 100644
--- a/native/csrc/ring/ring_engine_py.h
+++ b/native/csrc/ring/ring_engine_py.h
@@ -30,10 +30,10 @@ struct RingConfig {
     float    drain_flush_payload_ratio  = 0.5f;
     uint64_t drain_flush_entry_threshold = 0;
     uint64_t drain_flush_byte_threshold  = 0;
-    uint64_t drain_flush_timeout_us      = 100000;
+    uint64_t drain_flush_timeout_us      = 0;
     // Clone per-request slices
     bool     clone_slices               = false;
-    // Reserved for compatibility; StageConfig owns host queue limits.
+    // ClickHouse insert queue limits
     uint64_t insert_queue_max_bytes     = 4096ULL * 1024 * 1024;
     uint64_t insert_queue_max_items     = 65536;
 };
@@ -149,10 +149,15 @@ class RingEnginePy {
     //      HookPoint.forward (eager-only path).  Never called during
     //      CUDA-graph capture or replay.
 
-    uint64_t effective_capacity() const;
+    // Free bytes in the payload ring not currently reserved and not
+    // pending drain.  CPU-only read.
+    uint64_t available_capacity() const;
 
-    // Atomically check and reserve payload plus one task slot.
-    bool try_reserve_one(uint64_t nbytes);
+    // Per-hook reservation: claim `nbytes` of payload ring + 1 task entry
+    // for an upcoming producer kernel launch.  Used by the safety net
+    // when force_eager is on and the spec is dynamic-shape.  Advances
+    // cpu_payload_head/cpu_task_head atomically.
+    void reserve_one(uint64_t nbytes);
 
     // Synchronise the current CUDA stream + force drain to process all
     // outstanding entries.  Blocking; the Python binding releases the
diff --git a/native/csrc/ring/task_entry.h b/native/csrc/ring/task_entry.h
index 73199b543..f56ab0992 100644
--- a/native/csrc/ring/task_entry.h
+++ b/native/csrc/ring/task_entry.h
@@ -17,7 +17,7 @@ namespace ring {
 // Sentinel value for ready_seq -- indicates slot has not been published yet.
 //
 // Publish protocol:
-//   producer: write fields -> __threadfence_system() -> publish ready_seq
+//   producer: write all data fields -> __threadfence() -> write ready_seq = seq_no
 //   consumer: poll until __atomic_load_n(ready_seq) == expected -> read fields
 // ---------------------------------------------------------------------------
 static constexpr uint64_t READY_SEQ_SENTINEL = ~uint64_t(0);
diff --git a/native/csrc/ring/task_ring.cuh b/native/csrc/ring/task_ring.cuh
index 4897eea05..9e6978817 100644
--- a/native/csrc/ring/task_ring.cuh
+++ b/native/csrc/ring/task_ring.cuh
@@ -12,7 +12,7 @@
 //
 // Publish protocol (producer):
 //   1. Write all TaskEntry data fields at slot (head % capacity).
-//   2. __threadfence_system() -- makes data visible to the CPU consumer.
+//   2. __threadfence() -- ensures data is visible before ready_seq.
 //   3. Write ready_seq = head  (the slot's logical sequence number).
 //   4. Increment head.
 //
@@ -73,7 +73,7 @@ inline void task_ring_init(TaskEntry* d_entries, uint64_t capacity,
 // task_publish -- write a TaskEntry and publish it to the consumer.
 //
 // Copies all non-ready_seq fields from `src` into the slot at `seq_no %
-// capacity`, issues a system-scope fence to enforce write ordering, then writes
+// capacity`, issues a __threadfence() to enforce write ordering, then writes
 // ready_seq = seq_no.
 // ---------------------------------------------------------------------------
 __device__ inline void task_publish(
@@ -93,7 +93,7 @@ __device__ inline void task_publish(
     slot.payload_len2       = src.payload_len2;
 
     // Release fence: all stores above must be visible before ready_seq.
-    __threadfence_system();
+    __threadfence();
 
     // Publish: consumer spins until it sees this value.
     *reinterpret_cast<volatile uint64_t*>(&slot.ready_seq) = seq_no;
diff --git a/src/dmi/engine.py b/src/dmi/engine.py
index 42aa90d87..0405063dd 100644
--- a/src/dmi/engine.py
+++ b/src/dmi/engine.py
@@ -9,7 +9,7 @@
 from .config import MonitoringConfig
 
 
-DEFAULT_DRAIN_FLUSH_TIMEOUT_US = 100_000
+DEFAULT_DRAIN_FLUSH_TIMEOUT_US = 0
 
 
 def _native_module() -> Any:
@@ -280,13 +280,6 @@ def next_auto_group_id(self) -> int:
     def close(self) -> None:
         """Tear down backend resources."""
 
-        first_error: Optional[BaseException] = None
-
-        def capture_error(exc: BaseException) -> None:
-            nonlocal first_error
-            if first_error is None:
-                first_error = exc
-
         if self._ring_transport is not None:
             # Best-effort reset of the device-global native null flag.  This is
             # needed only after callers explicitly disabled capture; the normal
@@ -294,43 +287,30 @@ def capture_error(exc: BaseException) -> None:
             if not self.capture_enabled:
                 try:
                     self.set_capture_enabled(True)
-                except Exception as exc:
-                    capture_error(exc)
+                except Exception:
+                    pass
             try:
                 ring_engine = getattr(self, "_ring_engine", None)
                 if ring_engine is not None:
                     ring_engine.stop()
-            except Exception as exc:
-                capture_error(exc)
+            except Exception:
+                pass
             try:
                 _rt = _ring_module()
                 _rt.deactivate()
-            except Exception as exc:
-                capture_error(exc)
+            except Exception:
+                pass
             self._ring_transport = None
             self._ring_engine = None
 
         if self._host_engine is not None:
-            host_engine = self._host_engine
-            try:
-                host_engine.close_input()
-            except Exception as exc:
-                capture_error(exc)
-            try:
-                host_engine.stop()
-            except Exception as exc:
-                capture_error(exc)
             try:
-                raise_if_failed = getattr(host_engine, "raise_if_failed", None)
-                if raise_if_failed is not None:
-                    raise_if_failed()
-            except Exception as exc:
-                capture_error(exc)
+                self._host_engine.close_input()
+                self._host_engine.stop()
+            except Exception:
+                pass
             self._host_engine = None
 
-        if first_error is not None:
-            raise first_error
-
 
 # ---------------------------------------------------------------------------
 # Backend loader
diff --git a/src/dmi/hooks/point.py b/src/dmi/hooks/point.py
index 8c25cd376..e5c42a122 100644
--- a/src/dmi/hooks/point.py
+++ b/src/dmi/hooks/point.py
@@ -283,13 +283,13 @@ def forward(self, x: Tensor) -> Tensor:
             engine = transport._ring_engine
             if engine is not None:
                 nbytes = x_cont.nbytes
-                if engine.try_reserve_one(nbytes):
+                if nbytes <= engine.available_capacity():
+                    engine.reserve_one(nbytes)
                     dispatch_producer(ring_payload, x_cont, strip_t, strip_rb,
                                       self._ring_hook_type, self._ring_hook_id)
-                elif nbytes <= engine.effective_capacity():
+                elif nbytes <= engine.payload_cap():
                     engine.flush_and_wait()
-                    if not engine.try_reserve_one(nbytes):
-                        raise RuntimeError("Ring reservation failed after flush")
+                    engine.reserve_one(nbytes)
                     dispatch_producer(ring_payload, x_cont, strip_t, strip_rb,
                                       self._ring_hook_type, self._ring_hook_id)
                 else:
diff --git a/src/dmi/transport/ring.py b/src/dmi/transport/ring.py
index 44ddadc14..7b276a0ec 100644
--- a/src/dmi/transport/ring.py
+++ b/src/dmi/transport/ring.py
@@ -164,8 +164,8 @@ def __init__(self, ring_engine: Any) -> None:
 
         # When True, HookPoint.forward takes the runtime safety-net branch
         # instead of the fast path:
-        #   1. fits in current slack        -> try_reserve_one + ring
-        #   2. fits after flushing the ring -> flush + try_reserve_one + ring
+        #   1. fits in current slack       -> reserve_one + ring
+        #   2. fits after flushing the ring -> flush_and_wait + reserve_one + ring
         #   3. single tensor > ring        -> flush_and_wait + submit_cpu_direct
         # Owned by adaptor_base.before_forward (per-batch reassignment based
         # on prepare_step result and dynamic-spec presence).  Dispatch
diff --git a/tests/native/ring/Makefile b/tests/native/ring/Makefile
index 2c2b863f9..1a1b068b0 100644
--- a/tests/native/ring/Makefile
+++ b/tests/native/ring/Makefile
@@ -54,7 +54,6 @@ NVCC_FLAGS := \
 
 PRODUCER_CU  := ../../../native/csrc/ring/producer.cu
 DRAIN_SRC    := ../../../native/csrc/ring/drain_thread.cpp
-P2P_SRC      := ../../../native/csrc/ring/p2p_thread.cpp
 
 TARGETS := $(BUILD)/test_rings $(BUILD)/test_producer \
            $(BUILD)/test_ring_engine $(BUILD)/test_null_mode
@@ -77,7 +76,7 @@ $(BUILD)/test_rings: test_rings.cu $(CUDA_CONFIG_STAMP) | $(BUILD)
 $(BUILD)/test_producer: test_producer.cu $(PRODUCER_CU) $(CUDA_CONFIG_STAMP) | $(BUILD)
 	$(CUDA_NVCC) $(NVCC_FLAGS) $(filter-out $(CUDA_CONFIG_STAMP),$^) -o $@
 
-$(BUILD)/test_ring_engine: test_ring_engine.cu $(PRODUCER_CU) $(DRAIN_SRC) $(P2P_SRC) $(CUDA_CONFIG_STAMP) | $(BUILD)
+$(BUILD)/test_ring_engine: test_ring_engine.cu $(PRODUCER_CU) $(DRAIN_SRC) $(CUDA_CONFIG_STAMP) | $(BUILD)
 	$(CUDA_NVCC) $(NVCC_FLAGS) $(TORCH_INCLUDE_FLAGS) $(TORCH_LIB_FLAGS) \
 		$(TORCH_RPATH_FLAGS) -ltorch -ltorch_cpu -lc10 -lpthread \
 		$(filter-out $(CUDA_CONFIG_STAMP),$^) -o $@
diff --git a/tests/native/ring/test_producer.cu b/tests/native/ring/test_producer.cu
index df9056913..59012d78e 100644
--- a/tests/native/ring/test_producer.cu
+++ b/tests/native/ring/test_producer.cu
@@ -6,7 +6,6 @@
 #include <cuda_runtime.h>
 
 #include <algorithm>
-#include <chrono>
 #include <cstdint>
 #include <cstdio>
 #include <cstdlib>
@@ -277,31 +276,6 @@ static void test_serialized_static_launches() {
     }
 }
 
-static void test_host_observes_complete_published_descriptor() {
-    banner("host observes complete published descriptor without device sync");
-    ring::AllocatedRing allocated(make_config());
-    allocated.init();
-    ring::RingState& state = allocated.state();
-    const std::vector<uint8_t> source = pattern(333, 91);
-    uint8_t* device = upload(source);
-    cudaStream_t stream{};
-    CUDA_CHECK(cudaStreamCreateWithFlags(&stream, cudaStreamNonBlocking));
-
-    ring::launch_producer_static(state, device, source.size(), 5, stream);
-    const auto deadline = std::chrono::steady_clock::now() +
-                          std::chrono::seconds(5);
-    while (!ring::task_cpu_ready(state.task_entries, state.task_cap, 0) &&
-           std::chrono::steady_clock::now() < deadline) {
-    }
-
-    EXPECT(ring::task_cpu_ready(state.task_entries, state.task_cap, 0));
-    expect_entry(state, 0, source.size());
-
-    CUDA_CHECK(cudaStreamSynchronize(stream));
-    CUDA_CHECK(cudaStreamDestroy(stream));
-    CUDA_CHECK(cudaFree(device));
-}
-
 int main() {
     setbuf(stdout, nullptr);
     ring::set_ring_null_mode(false);
@@ -314,7 +288,6 @@ int main() {
     test_prefix_bounds();
     test_chunked_packed_copy();
     test_serialized_static_launches();
-    test_host_observes_complete_published_descriptor();
 
     std::printf("Results: %d passed, %d failed\n", g_pass, g_fail);
     return g_fail == 0 ? 0 : 1;
diff --git a/tests/native/ring/test_ring_engine.cu b/tests/native/ring/test_ring_engine.cu
index ee2535d27..806a85485 100644
--- a/tests/native/ring/test_ring_engine.cu
+++ b/tests/native/ring/test_ring_engine.cu
@@ -1,22 +1,17 @@
 // CUDA integration tests for producer -> drain -> pinned-staging delivery.
 
 #include "ring/drain_thread.h"
-#include "ring/p2p_thread.h"
 #include "ring/pinned_staging.h"
 #include "ring/producer.cuh"
 #include "ring/ring_alloc.h"
 
 #include <cuda_runtime.h>
 
-#include <atomic>
-#include <chrono>
 #include <cstdint>
 #include <cstdio>
 #include <cstdlib>
 #include <cstring>
 #include <memory>
-#include <stdexcept>
-#include <thread>
 #include <vector>
 
 static int g_pass = 0;
@@ -153,11 +148,10 @@ static void test_static_force_flush() {
     EXPECT(task.alloc_bytes == reserved);
     EXPECT(task.data_len1 + task.data_len2 == source.size());
     EXPECT(task_bytes(task) == source);
-    const auto capacity = harness.drain->capacity_snapshot();
-    EXPECT(capacity.task_head == 1);
-    EXPECT(capacity.task_tail_committed == 1);
-    EXPECT(capacity.payload_head == reserved);
-    EXPECT(capacity.payload_tail_committed == reserved);
+    EXPECT(harness.drain->cpu_task_head() == 1);
+    EXPECT(harness.drain->cpu_task_tail_committed() == 1);
+    EXPECT(harness.drain->cpu_payload_head() == reserved);
+    EXPECT(harness.drain->cpu_payload_tail_committed() == reserved);
 
     harness.release(task);
     CUDA_CHECK(cudaFree(device));
@@ -186,7 +180,7 @@ static void test_prefix_force_flush() {
     EXPECT(task.tensor_total_bytes == actual);
     EXPECT(task.alloc_bytes == actual);
     EXPECT(task_bytes(task) == expected);
-    EXPECT(harness.drain->capacity_snapshot().payload_tail_committed == actual);
+    EXPECT(harness.drain->cpu_payload_tail_committed() == actual);
 
     harness.release(task);
     CUDA_CHECK(cudaFree(device_count));
@@ -220,9 +214,8 @@ static void test_repeated_wrap_delivery() {
 
     EXPECT(*harness.allocated.state().task_head == 4);
     EXPECT(*harness.allocated.state().payload_head == 320);
-    const auto capacity_snapshot = harness.drain->capacity_snapshot();
-    EXPECT(capacity_snapshot.task_tail_committed == 4);
-    EXPECT(capacity_snapshot.payload_tail_committed == 320);
+    EXPECT(harness.drain->cpu_task_tail_committed() == 4);
+    EXPECT(harness.drain->cpu_payload_tail_committed() == 320);
     EXPECT(harness.staging.head() == 320);
     EXPECT(harness.staging.tail() == 320);
 }
@@ -241,123 +234,11 @@ static void test_zero_byte_delivery() {
     EXPECT(task.data_len1 == 0);
     EXPECT(task.data_ptr2 == nullptr);
     EXPECT(task.data_len2 == 0);
-    const auto capacity = harness.drain->capacity_snapshot();
-    EXPECT(capacity.task_tail_committed == 1);
-    EXPECT(capacity.payload_tail_committed == 0);
+    EXPECT(harness.drain->cpu_task_tail_committed() == 1);
+    EXPECT(harness.drain->cpu_payload_tail_committed() == 0);
     harness.release(task);
 }
 
-static void test_capacity_snapshot_is_consistent_during_reserve() {
-    banner("capacity snapshot is consistent during reserve");
-    DrainHarness harness(make_config(1ULL << 24));
-    constexpr uint64_t reservations = 10000;
-    constexpr uint64_t bytes_per_reservation = 64;
-    std::atomic<bool> done{false};
-
-    std::thread writer([&] {
-        for (uint64_t i = 0; i < reservations; ++i) {
-            harness.drain->reserve(bytes_per_reservation, 1);
-        }
-        done.store(true, std::memory_order_release);
-    });
-
-    while (!done.load(std::memory_order_acquire)) {
-        const auto snapshot = harness.drain->capacity_snapshot();
-        EXPECT(snapshot.payload_head ==
-               snapshot.task_head * bytes_per_reservation);
-    }
-    writer.join();
-
-    const auto snapshot = harness.drain->capacity_snapshot();
-    EXPECT(snapshot.payload_head == reservations * bytes_per_reservation);
-    EXPECT(snapshot.task_head == reservations);
-    EXPECT(snapshot.payload_tail_committed == 0);
-    EXPECT(snapshot.task_tail_committed == 0);
-}
-
-static void test_atomic_reservation_checks_effective_and_task_capacity() {
-    banner("atomic reservation checks effective and task capacity");
-    DrainHarness harness(make_config());
-
-    EXPECT(!harness.drain->try_reserve(257, 1, 256, 16));
-    for (uint32_t i = 0; i < 16; ++i) {
-        EXPECT(harness.drain->try_reserve(0, 1, 256, 16));
-    }
-    EXPECT(!harness.drain->try_reserve(0, 1, 256, 16));
-}
-
-static void test_force_flush_surfaces_invalid_d2h_source() {
-    banner("force flush surfaces invalid D2H source");
-    DrainHarness harness(make_config());
-    ring::RingState& state = harness.allocated.state();
-    uint8_t* payload_buf = state.payload_buf;
-    state.payload_buf = reinterpret_cast<uint8_t*>(1);
-
-    ring::TaskEntry& entry = state.task_entries[0];
-    entry.tensor_total_bytes = 64;
-    entry.payload_off1 = 0;
-    entry.payload_len1 = 64;
-    entry.payload_off2 = 0;
-    entry.payload_len2 = 0;
-    __atomic_store_n(&entry.ready_seq, uint64_t{0}, __ATOMIC_RELEASE);
-    harness.drain->reserve(64, 1);
-
-    bool threw = false;
-    try {
-        harness.drain->force_flush_and_wait();
-    } catch (const std::runtime_error&) {
-        threw = true;
-    }
-    EXPECT(threw);
-
-    state.payload_buf = payload_buf;
-}
-
-static void throwing_submit(
-    const std::string&, int32_t, const std::string&, const std::string&,
-    int32_t, int32_t, int32_t, at::Tensor)
-{
-    throw std::runtime_error("host submission failed");
-}
-
-static void test_p2p_submission_failure_reaches_drain_owner() {
-    banner("P2P submission failure reaches drain owner");
-    DrainHarness harness(make_config());
-    ring_py::TensorMetaFifo fifo;
-    auto* context = new ring_py::StepContext();
-    context->model_id = "test";
-    context->requests.push_back({"request", 0, 1, 0, 0});
-    std::vector<ring_py::TensorMeta> metas(1);
-    metas[0].hook_type = ring_py::HOOK_TYPE_TOKEN_IDS;
-    metas[0].shape = {1, 1};
-    metas[0].dtype = static_cast<int>(at::kByte);
-    metas[0].last_in_step = true;
-    fifo.push_step(context, metas);
-
-    ring::P2PThread p2p(*harness.drain, fifo, harness.cfg, throwing_submit);
-    p2p.start();
-    harness.drain->submit_cpu_direct(
-        at::zeros({1, 1}, at::TensorOptions().dtype(at::kByte)), 1);
-
-    const auto deadline = std::chrono::steady_clock::now() +
-                          std::chrono::seconds(5);
-    while (!harness.drain->has_failed() &&
-           std::chrono::steady_clock::now() < deadline) {
-        std::this_thread::yield();
-    }
-    EXPECT(harness.drain->has_failed());
-    bool threw = false;
-    try {
-        harness.drain->rethrow_if_failed();
-    } catch (const std::runtime_error&) {
-        threw = true;
-    }
-    EXPECT(threw);
-
-    harness.drain->signal_p2p_stop();
-    p2p.stop();
-}
-
 int main() {
     setbuf(stdout, nullptr);
     ring::set_ring_null_mode(false);
@@ -368,10 +249,6 @@ int main() {
     test_prefix_force_flush();
     test_repeated_wrap_delivery();
     test_zero_byte_delivery();
-    test_capacity_snapshot_is_consistent_during_reserve();
-    test_atomic_reservation_checks_effective_and_task_capacity();
-    test_force_flush_surfaces_invalid_d2h_source();
-    test_p2p_submission_failure_reaches_drain_owner();
 
     std::printf("Results: %d passed, %d failed\n", g_pass, g_fail);
     return g_fail == 0 ? 0 : 1;
diff --git a/tests/test_engine_runtime_api.py b/tests/test_engine_runtime_api.py
index ff7e72e91..108fe7cf1 100644
--- a/tests/test_engine_runtime_api.py
+++ b/tests/test_engine_runtime_api.py
@@ -10,17 +10,15 @@
 
 import pytest
 
-import dmi.engine as engine_module
 from dmi.engine import MonitoringEngine, RingCapacities
 
 pytestmark = pytest.mark.cpu
 
 
 class _FakeRingEngine:
-    def __init__(self, transport=None, *, fail_null_mode=False, fail_stop=False):
+    def __init__(self, transport=None, *, fail_null_mode=False):
         self.transport = transport
         self.fail_null_mode = fail_null_mode
-        self.fail_stop = fail_stop
         self.null_mode_calls = []
         self.stop_calls = 0
         self.init_calls = 0
@@ -57,12 +55,10 @@ def start(self):
 
     def stop(self):
         self.stop_calls += 1
-        if self.fail_stop:
-            raise RuntimeError("ring drain failed")
 
 
 def _engine_with_fake_ring(
-    *, null_offload=False, force_eager=False, fail_null_mode=False, fail_stop=False
+    *, null_offload=False, force_eager=False, fail_null_mode=False
 ):
     engine = MonitoringEngine(enable_ring_transport=False)
     transport = SimpleNamespace(
@@ -72,7 +68,6 @@ def _engine_with_fake_ring(
     ring_engine = _FakeRingEngine(
         transport,
         fail_null_mode=fail_null_mode,
-        fail_stop=fail_stop,
     )
     engine._ring_transport = transport
     engine._ring_engine = ring_engine
@@ -95,25 +90,6 @@ def test_ring_capacities_is_frozen_snapshot_with_effective_limit():
         capacities.payload_bytes = 1
 
 
-def test_default_ring_flushes_small_workloads_with_bounded_linger(monkeypatch):
-    class _RingConfig:
-        pass
-
-    monkeypatch.setattr(
-        engine_module,
-        "_native_module",
-        lambda: SimpleNamespace(RingConfig=_RingConfig),
-    )
-
-    config = MonitoringEngine._make_default_ring_config(
-        payload_mb=64,
-        pinned_mb=32,
-        task_entries=128,
-    )
-
-    assert config.drain_flush_timeout_us == 100_000
-
-
 def test_capture_toggle_changes_metadata_flag_after_native_transition():
     engine, transport, ring_engine = _engine_with_fake_ring(force_eager=True)
     assert engine.capture_enabled is True
@@ -167,75 +143,6 @@ def test_close_restores_device_global_null_mode_before_ring_stop():
     assert engine.capture_enabled is False
 
 
-def test_close_surfaces_host_worker_failure_after_cleanup(monkeypatch):
-    engine, _transport, ring_engine = _engine_with_fake_ring()
-    deactivated = []
-
-    class _FailingHostEngine:
-        def __init__(self):
-            self.calls = []
-
-        def close_input(self):
-            self.calls.append("close_input")
-
-        def stop(self):
-            self.calls.append("stop")
-
-        def raise_if_failed(self):
-            self.calls.append("raise_if_failed")
-            raise RuntimeError("clickhouse worker failed")
-
-    host_engine = _FailingHostEngine()
-    engine._host_engine = host_engine
-    fake_transport_module = ModuleType("dmi.transport.ring")
-    fake_transport_module.deactivate = lambda: deactivated.append(True)
-    monkeypatch.setitem(sys.modules, "dmi.transport.ring", fake_transport_module)
-
-    with pytest.raises(RuntimeError, match="clickhouse worker failed"):
-        engine.close()
-
-    assert ring_engine.stop_calls == 1
-    assert deactivated == [True]
-    assert host_engine.calls == ["close_input", "stop", "raise_if_failed"]
-    assert engine._ring_engine is None
-    assert engine._ring_transport is None
-    assert engine._host_engine is None
-
-
-def test_close_preserves_ring_failure_while_cleaning_up_host(monkeypatch):
-    engine, _transport, ring_engine = _engine_with_fake_ring(fail_stop=True)
-    deactivated = []
-
-    class _HostEngine:
-        def __init__(self):
-            self.calls = []
-
-        def close_input(self):
-            self.calls.append("close_input")
-
-        def stop(self):
-            self.calls.append("stop")
-
-        def raise_if_failed(self):
-            self.calls.append("raise_if_failed")
-
-    host_engine = _HostEngine()
-    engine._host_engine = host_engine
-    fake_transport_module = ModuleType("dmi.transport.ring")
-    fake_transport_module.deactivate = lambda: deactivated.append(True)
-    monkeypatch.setitem(sys.modules, "dmi.transport.ring", fake_transport_module)
-
-    with pytest.raises(RuntimeError, match="ring drain failed"):
-        engine.close()
-
-    assert ring_engine.stop_calls == 1
-    assert deactivated == [True]
-    assert host_engine.calls == ["close_input", "stop", "raise_if_failed"]
-    assert engine._ring_engine is None
-    assert engine._ring_transport is None
-    assert engine._host_engine is None
-
-
 def test_replacing_disabled_ring_restores_native_null_mode(monkeypatch):
     engine, _transport, old_ring = _engine_with_fake_ring(null_offload=True)
     new_ring = _FakeRingEngine()

From eb7ff7740efba2725d630a6377b106d89ca95728 Mon Sep 17 00:00:00 2001
From: "copilot-swe-agent[bot]" <198982749+Copilot@users.noreply.github.com>
Date: Tue, 25 Aug 2026 02:59:51 +0000
Subject: [PATCH 4/5] Reject non-finite timeout_s values in DMXHostEngine
 pybind lambdas

Co-authored-by: zaoxing <2923149+zaoxing@users.noreply.github.com>
---
 native/csrc/bindings.cpp | 14 +++++++++++++-
 1 file changed, 13 insertions(+), 1 deletion(-)

diff --git a/native/csrc/bindings.cpp b/native/csrc/bindings.cpp
index 7e33e5ed2..cf2b0524d 100644
--- a/native/csrc/bindings.cpp
+++ b/native/csrc/bindings.cpp
@@ -7,6 +7,7 @@
 #endif
 #include <pybind11/pybind11.h>
 #include <pybind11/stl.h>
+#include <cmath>
 namespace py = pybind11;
 
 #include "clickhouse_client.h"
@@ -232,6 +233,9 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
       .def("start", &DMXHostEngine::start)
       .def("wait_until_ready",
            [](DMXHostEngine& self, double timeout_s) {
+             if (!std::isfinite(timeout_s)) {
+               throw std::invalid_argument("timeout_s must be finite and non-negative");
+             }
              return self.wait_until_ready(DMXHostEngine::Duration(timeout_s));
            },
            py::arg("timeout_s"),
@@ -240,6 +244,9 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
       .def("stop",
            [](DMXHostEngine& self, bool graceful, std::optional<double> timeout_s) {
              if (timeout_s) {
+               if (!std::isfinite(*timeout_s)) {
+                 throw std::invalid_argument("timeout_s must be finite and non-negative");
+               }
                return self.stop(graceful, DMXHostEngine::Duration(*timeout_s));
              }
              return self.stop(graceful, std::nullopt);
@@ -251,7 +258,12 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) {
       .def("request_abort", &DMXHostEngine::request_abort)
       .def("join",
            [](DMXHostEngine& self, std::optional<double> timeout_s) {
-             if (timeout_s) return self.join(DMXHostEngine::Duration(*timeout_s));
+             if (timeout_s) {
+               if (!std::isfinite(*timeout_s)) {
+                 throw std::invalid_argument("timeout_s must be finite and non-negative");
+               }
+               return self.join(DMXHostEngine::Duration(*timeout_s));
+             }
              return self.join(std::nullopt);
            },
            py::arg("timeout_s") = std::optional<double>(),

From df0a6c667c864ac8a32f2948fd64215bae9db6f8 Mon Sep 17 00:00:00 2001
From: "copilot-swe-agent[bot]" <198982749+Copilot@users.noreply.github.com>
Date: Tue, 25 Aug 2026 03:16:28 +0000
Subject: [PATCH 5/5] Scope ServerTelemetrySampler process query to trial
 database/table

Co-authored-by: zaoxing <2923149+zaoxing@users.noreply.github.com>
---
 benchmarks/bench_clickhouse_host.py | 26 +++++++++++++++++++++++---
 1 file changed, 23 insertions(+), 3 deletions(-)

diff --git a/benchmarks/bench_clickhouse_host.py b/benchmarks/bench_clickhouse_host.py
index 4cee98a74..d091a8d52 100644
--- a/benchmarks/bench_clickhouse_host.py
+++ b/benchmarks/bench_clickhouse_host.py
@@ -241,7 +241,6 @@ def _client_metrics_as_dict(metrics: Any) -> dict[str, Any]:
 
 
 class ServerTelemetrySampler:
-    _PROCESS_QUERY = "SELECT count() FROM system.processes WHERE query_kind = 'Insert'"
     _METRICS_QUERY = """
         SELECT metric, value
         FROM system.metrics
@@ -259,11 +258,30 @@ class ServerTelemetrySampler:
                          'NetworkReceiveBytes', 'NetworkSendBytes')
     """
 
-    def __init__(self, client_factory: Callable[[], Any], interval_ms: int = 50):
+    def __init__(
+        self,
+        client_factory: Callable[[], Any],
+        interval_ms: int = 50,
+        database: str | None = None,
+        table: str | None = None,
+    ):
         if interval_ms <= 0:
             raise ValueError("interval_ms must be positive")
         self._client_factory = client_factory
         self._interval_seconds = interval_ms / 1000.0
+        if database is not None and table is not None:
+            qualified = f"{database}.{table}"
+            self._process_query = (
+                "SELECT count() FROM system.processes"
+                " WHERE query_kind = 'Insert'"
+                " AND has(tables, %(table)s)"
+            )
+            self._process_query_params: dict[str, Any] = {"table": qualified}
+        else:
+            self._process_query = (
+                "SELECT count() FROM system.processes WHERE query_kind = 'Insert'"
+            )
+            self._process_query_params = {}
         self._stop = threading.Event()
         self._thread: threading.Thread | None = None
         self._client: Any = None
@@ -307,7 +325,7 @@ def sample_once(self) -> None:
         try:
             if self._client is None:
                 self._client = self._client_factory()
-            active = int(self._client.execute(self._PROCESS_QUERY)[0][0])
+            active = int(self._client.execute(self._process_query, self._process_query_params)[0][0])
             metrics = self._client.execute(self._METRICS_QUERY)
             async_metrics = self._client.execute(self._ASYNC_METRICS_QUERY)
             with self._lock:
@@ -681,6 +699,8 @@ def run(config: BenchmarkConfig) -> dict[str, Any]:
             sampler = ServerTelemetrySampler(
                 lambda: _connect(config, min(2.0, config.socket_timeout_seconds)),
                 interval_ms=config.server_sample_interval_ms,
+                database=config.database,
+                table=config.table,
             )
         measurement = _submit_and_drain(
             _build_engine(config),