feat: add new vgwrdma service for cuObject compatible server

2026-08-16 12:16:14 +00:00 · 2026-08-04 08:48:17 -07:00
parent 5422bf8e8c
commit 62c6787a17
44 changed files with 5315 additions and 57 deletions
@@ -0,0 +1,38 @@
+name: vgwrdma build
+permissions: {}
+on: pull_request
+
+jobs:
+
+  vgwrdma:
+    name: Build vgwrdma (linux/amd64, cgo)
+    runs-on: ubuntu-latest
+    steps:
+
+    - name: Check out code
+      uses: actions/checkout@v7
+
+    - name: make vgwrdma-docker
+      run: make vgwrdma-docker
+
+  cuobjtest-gpu:
+    name: Build cuobjtest-gpu (linux/amd64, cgo)
+    runs-on: ubuntu-latest
+    steps:
+
+    - name: Check out code
+      uses: actions/checkout@v7
+
+    - name: make cuobjtest-gpu-docker
+      run: make cuobjtest-gpu-docker
+
+  cuobjtest-host:
+    name: Build cuobjtest-host (linux/amd64, cgo)
+    runs-on: ubuntu-latest
+    steps:
+
+    - name: Check out code
+      uses: actions/checkout@v7
+
+    - name: make cuobjtest-host-docker
+      run: make cuobjtest-host-docker
@@ -10,6 +10,12 @@
 .DS_Store
 cmd/versitygw/versitygw
 /versitygw
+/vgwrdma
+/cuobjtest
+/.cache/
+/rdma/libcuobjwrapper.a
+/rdma/libcuobjclientwrapper.a
+/rdma/libhostclientwrapper.a

 # Test binary, built with `go test -c`
 *.test
@@ -23,6 +23,37 @@ DCCMD=docker-compose
 DOCKERCOMPOSE=$(DCCMD) -f tests/docker-compose.yml --env-file .env.dev --project-directory .

 BIN=versitygw
+VGWRDMA_BIN=vgwrdma
+VGWRDMA_CMD=./cmd/vgwrdma
+VGWRDMA_BUILDER_DOCKERFILE ?= build/vgwrdma-builder/Dockerfile
+VGWRDMA_BUILDER_IMAGE ?= vgwrdma-builder:local
+VGWRDMA_BUILDER_PLATFORM ?= linux/amd64
+CUOBJTEST_BIN=cuobjtest
+CUOBJTEST_CMD=./cmd/cuobjtest
+CUOBJTEST_HOST_TAG ?= cuobjclient_host
+CUOBJTEST_GPU_BUILDER_DOCKERFILE ?= build/cuobjtest-gpu-builder/Dockerfile
+CUOBJTEST_GPU_BUILDER_IMAGE ?= cuobjtest-gpu-builder:local
+CUOBJTEST_GPU_BUILDER_PLATFORM ?= linux/amd64
+CUOBJTEST_HOST_BUILDER_DOCKERFILE ?= build/cuobjtest-host-builder/Dockerfile
+CUOBJTEST_HOST_BUILDER_IMAGE ?= cuobjtest-host-builder:local
+CUOBJTEST_HOST_BUILDER_PLATFORM ?= linux/amd64
+
+# RDMA build dependencies (Linux/amd64 + cgo)
+CUOBJ_LIB_DIR ?= /usr/local/cuda-13.3/targets/x86_64-linux/lib
+CUOBJ_SERVER_INC_DIR ?= /usr/include
+CUOBJ_CLIENT_INC_DIR ?= /usr/include
+CUOBJ_CUDA_INC_DIR ?= /usr/local/cuda/include
+CXX ?= g++
+AR ?= ar
+
+# Static archive for the cuObjServer C++ wrapper used by the RDMA gateway.
+VGWRDMA_WRAPPER_LIB=rdma/libcuobjwrapper.a
+VGWRDMA_RDMA_CGO_CFLAGS=-I$(CUOBJ_SERVER_INC_DIR)
+VGWRDMA_RDMA_CGO_LDFLAGS=-L$(CUOBJ_LIB_DIR) -Wl,-rpath,$(CUOBJ_LIB_DIR)
+CUOBJCLIENT_WRAPPER_LIB=rdma/libcuobjclientwrapper.a
+CUOBJCLIENT_CGO_CFLAGS=-I$(CUOBJ_CLIENT_INC_DIR) -I$(CUOBJ_CUDA_INC_DIR)
+CUOBJCLIENT_CGO_LDFLAGS=-L$(CUOBJ_LIB_DIR) -Wl,-rpath,$(CUOBJ_LIB_DIR)
+HOSTCLIENT_WRAPPER_LIB=rdma/libhostclientwrapper.a

 VERSION := $(shell if test -e VERSION; then cat VERSION; else git describe --abbrev=0 --tags HEAD; fi)
 BUILD := $(shell git rev-parse --short HEAD || echo release-rpm)
@@ -38,6 +69,80 @@ build: $(BIN)
 $(BIN):
 	$(GOBUILD) $(LDFLAGS) -o $(BIN) cmd/$(BIN)/*.go

+$(VGWRDMA_WRAPPER_LIB): cuwrapper/cuobjserver_wrapper.cpp cuwrapper/cuobjserver_wrapper.h
+	$(CXX) -c -fPIC \
+		-I$(CUOBJ_SERVER_INC_DIR) -Icuwrapper \
+		-o cuwrapper/cuobjserver_wrapper.o \
+		cuwrapper/cuobjserver_wrapper.cpp
+	$(AR) rcs $(VGWRDMA_WRAPPER_LIB) cuwrapper/cuobjserver_wrapper.o
+	rm -f cuwrapper/cuobjserver_wrapper.o
+
+$(CUOBJCLIENT_WRAPPER_LIB): cuwrapper/cuobjclient_wrapper.cpp cuwrapper/cuobjclient_wrapper.h
+	$(CXX) -c -fPIC \
+		-I$(CUOBJ_CLIENT_INC_DIR) -I$(CUOBJ_CUDA_INC_DIR) -Icuwrapper \
+		-o cuwrapper/cuobjclient_wrapper.o \
+		cuwrapper/cuobjclient_wrapper.cpp
+	$(AR) rcs $(CUOBJCLIENT_WRAPPER_LIB) cuwrapper/cuobjclient_wrapper.o
+	rm -f cuwrapper/cuobjclient_wrapper.o
+
+$(HOSTCLIENT_WRAPPER_LIB): cuwrapper/rdma_host_client_wrapper.cpp cuwrapper/rdma_host_client_wrapper.h
+	$(CXX) -c -fPIC \
+		-Icuwrapper \
+		-o cuwrapper/rdma_host_client_wrapper.o \
+		cuwrapper/rdma_host_client_wrapper.cpp
+	$(AR) rcs $(HOSTCLIENT_WRAPPER_LIB) cuwrapper/rdma_host_client_wrapper.o
+	rm -f cuwrapper/rdma_host_client_wrapper.o
+
+.PHONY: vgwrdma
+vgwrdma: $(VGWRDMA_WRAPPER_LIB)
+	CGO_ENABLED=1 \
+	CGO_CFLAGS="$(VGWRDMA_RDMA_CGO_CFLAGS)" \
+	CGO_LDFLAGS="$(VGWRDMA_RDMA_CGO_LDFLAGS)" \
+		$(GOBUILD) -buildvcs=false $(LDFLAGS) -o $(VGWRDMA_BIN) $(VGWRDMA_CMD)
+
+.PHONY: cuobjtest
+cuobjtest: cuobjtest-gpu
+
+.PHONY: cuobjtest-gpu
+cuobjtest-gpu: $(CUOBJCLIENT_WRAPPER_LIB)
+	CGO_ENABLED=1 \
+	CGO_CFLAGS="$(CUOBJCLIENT_CGO_CFLAGS)" \
+	CGO_LDFLAGS="$(CUOBJCLIENT_CGO_LDFLAGS)" \
+		$(GOBUILD) -buildvcs=false $(LDFLAGS) -o $(CUOBJTEST_BIN) $(CUOBJTEST_CMD)
+
+.PHONY: cuobjtest-host
+cuobjtest-host: $(HOSTCLIENT_WRAPPER_LIB)
+	CGO_ENABLED=1 \
+		$(GOBUILD) -buildvcs=false -tags $(CUOBJTEST_HOST_TAG) $(LDFLAGS) -o $(CUOBJTEST_BIN) $(CUOBJTEST_CMD)
+
+.PHONY: vgwrdma-builder-image
+vgwrdma-builder-image:
+	docker build --platform $(VGWRDMA_BUILDER_PLATFORM) -f $(VGWRDMA_BUILDER_DOCKERFILE) -t $(VGWRDMA_BUILDER_IMAGE) .
+
+.PHONY: vgwrdma-docker
+vgwrdma-docker: vgwrdma-builder-image
+	docker run --rm --platform $(VGWRDMA_BUILDER_PLATFORM) -v "$(CURDIR)":/workspace -w /workspace $(VGWRDMA_BUILDER_IMAGE)
+
+.PHONY: cuobjtest-gpu-builder-image
+cuobjtest-gpu-builder-image:
+	docker build --platform $(CUOBJTEST_GPU_BUILDER_PLATFORM) -f $(CUOBJTEST_GPU_BUILDER_DOCKERFILE) -t $(CUOBJTEST_GPU_BUILDER_IMAGE) .
+
+.PHONY: cuobjtest-gpu-docker
+cuobjtest-gpu-docker: cuobjtest-gpu-builder-image
+	docker run --rm --platform $(CUOBJTEST_GPU_BUILDER_PLATFORM) \
+		-v "$(CURDIR)":/workspace \
+		-w /workspace $(CUOBJTEST_GPU_BUILDER_IMAGE)
+
+.PHONY: cuobjtest-host-builder-image
+cuobjtest-host-builder-image:
+	docker build --platform $(CUOBJTEST_HOST_BUILDER_PLATFORM) -f $(CUOBJTEST_HOST_BUILDER_DOCKERFILE) -t $(CUOBJTEST_HOST_BUILDER_IMAGE) .
+
+.PHONY: cuobjtest-host-docker
+cuobjtest-host-docker: cuobjtest-host-builder-image
+	docker run --rm --platform $(CUOBJTEST_HOST_BUILDER_PLATFORM) \
+		-v "$(CURDIR)":/workspace \
+		-w /workspace $(CUOBJTEST_HOST_BUILDER_IMAGE)
+
 testbin:
 	$(GOBUILD) $(LDFLAGS) -o $(BIN) -cover -race cmd/$(BIN)/*.go

@@ -61,6 +166,11 @@ clean:
 .PHONY: cleanall
 cleanall: clean
 	rm -f $(BIN)
+	rm -f $(VGWRDMA_BIN)
+	rm -f $(CUOBJTEST_BIN)
+	rm -f $(VGWRDMA_WRAPPER_LIB)
+	rm -f $(CUOBJCLIENT_WRAPPER_LIB)
+	rm -f $(HOSTCLIENT_WRAPPER_LIB)
 	rm -f versitygw-*.tar
 	rm -f versitygw-*.tar.gz

@@ -25,6 +25,12 @@ Get more details about the new (optional) WebGUI management/explorer here: [http

 ![admin-explorer](https://github.com/user-attachments/assets/e99db171-2c72-4d0f-8c8d-480a56e1c8a1)

+### S3 RDMA
+VersityGW supports S3 over RDMA (Remote Direct Memory Access), enabling high-throughput, low-latency object transfers that bypass the kernel network stack. This is particularly useful for HPC and data-intensive workloads where network overhead is a bottleneck.
+See the [S3 RDMA](https://github.com/versity/versitygw/wiki/S3-RDMA) wiki page for setup and usage details.
+
+`vgwrdma` is a VersityGW-based service that exposes a standard S3 API accelerated with NVIDIA's cuObject (GPUDirect Storage for Objects) protocol.
+
 ### Static Website Hosting
 Serve S3 buckets as static websites with index documents, custom error pages, and routing rules.
 Enable a separate website endpoint with `--website :8090 --website-domain example.com` for virtual-host style routing (`blog.example.com` serves bucket `blog`, `example.com` serves bucket `example.com`).
@@ -0,0 +1,40 @@
+# syntax=docker/dockerfile:1.7
+
+ARG CUDA_TAG=13.3.0-devel-rockylinux9
+FROM --platform=linux/amd64 nvidia/cuda:${CUDA_TAG}
+
+ARG CUDA_TAG
+ARG CUOBJCLIENT_PKG_SUFFIX=
+
+# Install build prerequisites and cuObject client development packages.
+RUN set -eux; \
+    cuda_version="${CUDA_TAG%%-*}"; \
+    cuda_suffix="$(printf '%s' "$cuda_version" | awk -F. '{print $1"-"$2}')"; \
+    pkg_suffix="${CUOBJCLIENT_PKG_SUFFIX:-$cuda_suffix}"; \
+    dnf -y install 'dnf-command(config-manager)'; \
+    dnf config-manager --add-repo https://developer.download.nvidia.com/compute/cuda/repos/rhel9/x86_64/cuda-rhel9.repo; \
+    dnf -y makecache; \
+    dnf -y install \
+      libcuobjclient-${pkg_suffix} \
+      libcuobjclient-devel-${pkg_suffix} \
+      rdma-core-devel \
+      numactl-libs \
+      gcc \
+      gcc-c++ \
+      make \
+      golang \
+      git \
+      findutils \
+      which; \
+    dnf clean all
+
+WORKDIR /workspace
+
+# Match current package install layout discovered in the NVIDIA Rocky image.
+ENV CUOBJ_LIB_DIR=/usr/local/cuda/targets/x86_64-linux/lib
+ENV CUOBJ_CLIENT_INC_DIR=/usr/include
+ENV CUOBJ_CUDA_INC_DIR=/usr/local/cuda/include
+ENV CUOBJCLIENT_CGO_LDFLAGS="-L/usr/local/cuda/targets/x86_64-linux/lib -Wl,-rpath,/usr/local/cuda/targets/x86_64-linux/lib -Wl,-rpath-link,/usr/local/cuda/targets/x86_64-linux/lib"
+
+# Default command builds GPU-mode cuobjtest in a mounted workspace checkout.
+CMD ["bash", "-lc", "make cuobjtest-gpu CUOBJ_LIB_DIR=$CUOBJ_LIB_DIR CUOBJ_CLIENT_INC_DIR=$CUOBJ_CLIENT_INC_DIR CUOBJ_CUDA_INC_DIR=$CUOBJ_CUDA_INC_DIR CUOBJCLIENT_CGO_LDFLAGS=\"$CUOBJCLIENT_CGO_LDFLAGS\""]
@@ -0,0 +1,21 @@
+# syntax=docker/dockerfile:1.7
+
+FROM --platform=linux/amd64 rockylinux:9
+
+# Install build prerequisites for host-memory RDMA cuobjtest mode.
+RUN set -eux; \
+    dnf -y install \
+      rdma-core-devel \
+      gcc \
+      gcc-c++ \
+      make \
+      golang \
+      git \
+      findutils \
+      which; \
+    dnf clean all
+
+WORKDIR /workspace
+
+# Default command builds host-mode cuobjtest in a mounted workspace checkout.
+CMD ["bash", "-lc", "make cuobjtest-host"]
@@ -0,0 +1,33 @@
+# syntax=docker/dockerfile:1.7
+
+ARG CUDA_TAG=13.3.0-devel-rockylinux9
+FROM --platform=linux/amd64 nvidia/cuda:${CUDA_TAG}
+
+# Install build prerequisites and cuObject server development packages.
+RUN set -eux; \
+    dnf -y install 'dnf-command(config-manager)'; \
+    dnf config-manager --add-repo https://developer.download.nvidia.com/compute/cuda/repos/rhel9/x86_64/cuda-rhel9.repo; \
+    dnf -y makecache; \
+    dnf -y install \
+      cuobjserver \
+      libcuobjserver-devel \
+      rdma-core-devel \
+      numactl-libs \
+      gcc \
+      gcc-c++ \
+      make \
+      golang \
+      git \
+      findutils \
+      which; \
+    dnf clean all
+
+WORKDIR /workspace
+
+# Match current package install layout discovered in the NVIDIA Rocky image.
+ENV CUOBJ_LIB_DIR=/usr/lib64
+ENV CUOBJ_SERVER_INC_DIR=/usr/include
+ENV VGWRDMA_RDMA_CGO_LDFLAGS="-L/usr/lib64 -Wl,-rpath,/usr/lib64 -Wl,-rpath-link,/usr/lib64"
+
+# Default command builds vgwrdma in a mounted workspace checkout.
+CMD ["bash", "-lc", "make vgwrdma CUOBJ_LIB_DIR=$CUOBJ_LIB_DIR CUOBJ_SERVER_INC_DIR=$CUOBJ_SERVER_INC_DIR VGWRDMA_RDMA_CGO_LDFLAGS=\"$VGWRDMA_RDMA_CGO_LDFLAGS\""]
@@ -0,0 +1,519 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+package main
+
+import (
+	"bytes"
+	"context"
+	crand "crypto/rand"
+	"errors"
+	"flag"
+	"fmt"
+	"hash/crc32"
+	"io"
+	"math"
+	"os"
+	"sort"
+	"strconv"
+	"strings"
+	"time"
+
+	"github.com/aws/aws-sdk-go-v2/aws"
+	"github.com/aws/aws-sdk-go-v2/credentials"
+	s3lib "github.com/aws/aws-sdk-go-v2/service/s3"
+	s3types "github.com/aws/aws-sdk-go-v2/service/s3/types"
+
+	"github.com/versity/versitygw/cuobjclient"
+)
+
+var (
+	endpoint   = flag.String("endpoint", "http://localhost:7070", "cuserver S3 endpoint URL")
+	bucket     = flag.String("bucket", "cuobjtest", "S3 bucket (created if absent)")
+	key        = flag.String("key", "cuobjtest-object", "S3 object key prefix")
+	sizeStr    = flag.String("size", "64MiB", "Transfer size per iteration: e.g. 4MiB, 256MiB, 1GiB")
+	access     = flag.String("access", "admin", "S3 access key")
+	secret     = flag.String("secret", "", "S3 secret key (default: $AWS_SECRET_ACCESS_KEY); prefer the environment variable over this flag, which is visible in shell history and process listings")
+	region     = flag.String("region", "us-east-1", "S3 region")
+	iterations = flag.Int("n", 3, "Number of benchmark iterations")
+	putOnly    = flag.Bool("put-only", false, "Run PUT only (skip GET and checksum verification)")
+	getOnly    = flag.Bool("get-only", false, "Run GET only (assumes object already exists)")
+	stdS3      = flag.Bool("std-s3", false, "Use standard S3 PUT/GET (no cuObject RDMA transfer)")
+)
+
+func main() {
+	flag.Parse()
+
+	if *secret == "" {
+		*secret = os.Getenv("AWS_SECRET_ACCESS_KEY")
+	}
+	if *secret == "" {
+		fmt.Fprintln(os.Stderr, "cuobjtest: -secret or AWS_SECRET_ACCESS_KEY is required")
+		flag.Usage()
+		os.Exit(1)
+	}
+	if *putOnly && *getOnly {
+		fatalf("cuobjtest: -put-only and -get-only are mutually exclusive")
+	}
+	if *iterations < 1 {
+		fatalf("cuobjtest: -n must be >= 1")
+	}
+
+	size, err := parseSize(*sizeStr)
+	if err != nil {
+		fatalf("cuobjtest: -size: %v", err)
+	}
+	if size <= 0 || size > cuobjclient.MaxTransferSize {
+		fatalf("cuobjtest: -size must be between 1 B and %d bytes", cuobjclient.MaxTransferSize)
+	}
+
+	if err := runBenchmark(size); err != nil {
+		fatalf("cuobjtest: %v", err)
+	}
+}
+
+type iterationResult struct {
+	index  int
+	putDur time.Duration
+	getDur time.Duration
+	status string
+	err    error
+}
+
+type benchmarkStats struct {
+	putDurs  []time.Duration
+	getDurs  []time.Duration
+	mismatch int
+}
+
+func runBenchmark(size int) error {
+	base := newS3Client(*endpoint, *access, *secret, *region)
+	if err := ensureBucket(base, *bucket); err != nil {
+		return err
+	}
+
+	printRunHeader(size)
+
+	stats := benchmarkStats{}
+	var results []iterationResult
+	var benchmarkDur time.Duration
+	if *stdS3 {
+		results, benchmarkDur = runConcurrentStdS3(base, size)
+	} else {
+		results, benchmarkDur = runConcurrent(base, size)
+	}
+	sort.Slice(results, func(i, j int) bool {
+		return results[i].index < results[j].index
+	})
+	for _, r := range results {
+		if r.err != nil {
+			return fmt.Errorf("iter %d: %w", r.index, r.err)
+		}
+		accumulateStats(&stats, r)
+		printIterationResult(size, r)
+	}
+
+	return printSummary(stats, benchmarkDur, size)
+}
+
+func printRunHeader(size int) {
+	mode := runModeLabel()
+	if *stdS3 {
+		mode = "standard-s3"
+	}
+	fmt.Printf("\ncuobjtest    endpoint=%-28s  size=%s  iterations=%d\n", *endpoint, formatSize(size), *iterations)
+	fmt.Printf("            mode=%s\n\n", mode)
+	if *putOnly {
+		fmt.Printf("%-4s  %-14s %-10s\n", "#", "PUT latency", "PUT GB/s")
+	} else if *getOnly {
+		fmt.Printf("%-4s  %-14s %-10s\n", "#", "GET latency", "GET GB/s")
+	} else {
+		fmt.Printf("%-4s  %-14s %-10s   %-14s %-10s   %s\n",
+			"#", "PUT latency", "PUT GB/s", "GET latency", "GET GB/s", "checksum")
+	}
+	fmt.Printf("%s\n", strings.Repeat("-", 70))
+}
+
+func accumulateStats(stats *benchmarkStats, r iterationResult) {
+	if !*getOnly {
+		stats.putDurs = append(stats.putDurs, r.putDur)
+	}
+	if !*putOnly {
+		stats.getDurs = append(stats.getDurs, r.getDur)
+	}
+	if r.status != "OK" {
+		stats.mismatch++
+	}
+}
+
+func printSummary(stats benchmarkStats, benchmarkDur time.Duration, size int) error {
+
+	fmt.Printf("%s\n", strings.Repeat("-", 70))
+	putCount := len(stats.putDurs)
+	getCount := len(stats.getDurs)
+	putBytes := int64(putCount * size)
+	getBytes := int64(getCount * size)
+	totalBytes := putBytes + getBytes
+	// PUT/GET throughput is measured against the time actually spent in
+	// PUT/GET calls, not the combined benchmarkDur (which also includes
+	// the other operation and buffer preparation).
+	putDur := sumDurs(stats.putDurs)
+	getDur := sumDurs(stats.getDurs)
+
+	fmt.Printf("OVERALL elapsed=%s", fmtDur(benchmarkDur))
+	if putCount > 0 {
+		fmt.Printf("  PUT agg=%s (%d ops, %.2f ops/s)",
+			fmt.Sprintf("%.3f GB/s", float64(putBytes)/putDur.Seconds()/1e9),
+			putCount,
+			float64(putCount)/putDur.Seconds())
+	}
+	if getCount > 0 {
+		fmt.Printf("  GET agg=%s (%d ops, %.2f ops/s)",
+			fmt.Sprintf("%.3f GB/s", float64(getBytes)/getDur.Seconds()/1e9),
+			getCount,
+			float64(getCount)/getDur.Seconds())
+	}
+	if putCount > 0 && getCount > 0 {
+		fmt.Printf("  COMBINED=%s",
+			fmt.Sprintf("%.3f GB/s", float64(totalBytes)/benchmarkDur.Seconds()/1e9))
+	}
+	fmt.Printf("\n")
+
+	if *iterations > 1 {
+		if !*getOnly {
+			pa, plo, phi := durStats(stats.putDurs)
+			fmt.Printf("PUT  avg=%-10s  min=%-10s  max=%-10s  avg %s\n",
+				fmtDur(pa), fmtDur(plo), fmtDur(phi), fmtGBps(size, pa))
+		}
+		if !*putOnly {
+			ga, glo, ghi := durStats(stats.getDurs)
+			fmt.Printf("GET  avg=%-10s  min=%-10s  max=%-10s  avg %s\n",
+				fmtDur(ga), fmtDur(glo), fmtDur(ghi), fmtGBps(size, ga))
+		}
+	}
+
+	if *getOnly {
+		fmt.Printf("\nGET-only run complete.\n")
+		return nil
+	}
+	if *putOnly {
+		fmt.Printf("\nPUT-only run complete.\n")
+		return nil
+	}
+	if stats.mismatch > 0 {
+		return fmt.Errorf("%d/%d iterations had checksum mismatches", stats.mismatch, *iterations)
+	}
+	fmt.Printf("\nAll checksums OK.\n")
+	return nil
+}
+
+func runConcurrent(base *s3lib.Client, size int) ([]iterationResult, time.Duration) {
+	session, err := cuobjclient.NewSession(size)
+	if err != nil {
+		return []iterationResult{{index: 1, err: fmt.Errorf("init cuObj session: %w", err)}}, 0
+	}
+	defer session.Close()
+
+	putHost := make([]byte, size)
+	getHost := make([]byte, size)
+	out := make([]iterationResult, 0, *iterations)
+
+	benchmarkStart := time.Now()
+	for i := 1; i <= *iterations; i++ {
+		res := iterationResult{index: i, status: "OK"}
+		iterKey := *key
+
+		if !*getOnly {
+			if _, err := crand.Read(putHost); err != nil {
+				res.err = fmt.Errorf("fill PUT buffer: %w", err)
+				out = append(out, res)
+				continue
+			}
+			putCRC := crc32.ChecksumIEEE(putHost)
+
+			putStart := time.Now()
+			err := session.Upload(base, *bucket, iterKey, putHost)
+			res.putDur = time.Since(putStart)
+			if err != nil {
+				res.err = fmt.Errorf("PUT: %w", err)
+				out = append(out, res)
+				continue
+			}
+
+			if *putOnly {
+				out = append(out, res)
+				continue
+			}
+
+			getStart := time.Now()
+			err = session.Download(base, *bucket, iterKey, getHost)
+			res.getDur = time.Since(getStart)
+			if err != nil {
+				res.err = fmt.Errorf("GET: %w", err)
+				out = append(out, res)
+				continue
+			}
+			getCRC := crc32.ChecksumIEEE(getHost)
+			if getCRC != putCRC {
+				res.status = fmt.Sprintf("MISMATCH put=%08x get=%08x", putCRC, getCRC)
+			}
+			out = append(out, res)
+			continue
+		}
+
+		getStart := time.Now()
+		err := session.Download(base, *bucket, iterKey, getHost)
+		res.getDur = time.Since(getStart)
+		if err != nil {
+			res.err = fmt.Errorf("GET: %w", err)
+		}
+		out = append(out, res)
+	}
+
+	return out, time.Since(benchmarkStart)
+}
+
+// runConcurrentStdS3 exercises plain S3 PUT/GET (request body carries the
+// actual data, no RDMA headers) so it can be benchmarked against the
+// cuObject RDMA path in runConcurrent.
+func runConcurrentStdS3(base *s3lib.Client, size int) ([]iterationResult, time.Duration) {
+	putHost := make([]byte, size)
+	getHost := make([]byte, size)
+	out := make([]iterationResult, 0, *iterations)
+
+	benchmarkStart := time.Now()
+	for i := 1; i <= *iterations; i++ {
+		res := iterationResult{index: i, status: "OK"}
+		iterKey := *key
+
+		if !*getOnly {
+			if _, err := crand.Read(putHost); err != nil {
+				res.err = fmt.Errorf("fill PUT buffer: %w", err)
+				out = append(out, res)
+				continue
+			}
+			putCRC := crc32.ChecksumIEEE(putHost)
+
+			putStart := time.Now()
+			err := s3PutObject(base, *bucket, iterKey, putHost)
+			res.putDur = time.Since(putStart)
+			if err != nil {
+				res.err = fmt.Errorf("PUT: %w", err)
+				out = append(out, res)
+				continue
+			}
+
+			if *putOnly {
+				out = append(out, res)
+				continue
+			}
+
+			getStart := time.Now()
+			err = s3GetObject(base, *bucket, iterKey, getHost)
+			res.getDur = time.Since(getStart)
+			if err != nil {
+				res.err = fmt.Errorf("GET: %w", err)
+				out = append(out, res)
+				continue
+			}
+			getCRC := crc32.ChecksumIEEE(getHost)
+			if getCRC != putCRC {
+				res.status = fmt.Sprintf("MISMATCH put=%08x get=%08x", putCRC, getCRC)
+			}
+			out = append(out, res)
+			continue
+		}
+
+		getStart := time.Now()
+		err := s3GetObject(base, *bucket, iterKey, getHost)
+		res.getDur = time.Since(getStart)
+		if err != nil {
+			res.err = fmt.Errorf("GET: %w", err)
+		}
+		out = append(out, res)
+	}
+
+	return out, time.Since(benchmarkStart)
+}
+
+// s3PutObject issues a standard S3 PUT carrying data in the request body.
+func s3PutObject(base *s3lib.Client, bucket, key string, data []byte) error {
+	_, err := base.PutObject(context.Background(), &s3lib.PutObjectInput{
+		Bucket:        aws.String(bucket),
+		Key:           aws.String(key),
+		Body:          bytes.NewReader(data),
+		ContentLength: aws.Int64(int64(len(data))),
+	})
+	return err
+}
+
+// s3GetObject issues a standard S3 GET and reads the full body into dst.
+func s3GetObject(base *s3lib.Client, bucket, key string, dst []byte) error {
+	out, err := base.GetObject(context.Background(), &s3lib.GetObjectInput{
+		Bucket: aws.String(bucket),
+		Key:    aws.String(key),
+	})
+	if err != nil {
+		return err
+	}
+	defer out.Body.Close()
+	// io.ReadFull only guarantees len(dst) bytes were read; it doesn't
+	// notice a larger object whose extra bytes are simply left unread.
+	// Compare against the reported object size so a too-small dst is
+	// caught instead of silently benchmarking a truncated prefix.
+	if out.ContentLength != nil && *out.ContentLength != int64(len(dst)) {
+		return fmt.Errorf("object size %d does not match expected size %d", *out.ContentLength, len(dst))
+	}
+	_, err = io.ReadFull(out.Body, dst)
+	return err
+}
+
+func newS3Client(endpoint, access, secret, region string) *s3lib.Client {
+	return s3lib.New(s3lib.Options{
+		BaseEndpoint: aws.String(endpoint),
+		Region:       region,
+		Credentials:  aws.NewCredentialsCache(credentials.NewStaticCredentialsProvider(access, secret, "")),
+		UsePathStyle: true,
+	})
+}
+
+func ensureBucket(c *s3lib.Client, bucket string) error {
+	_, err := c.HeadBucket(context.Background(), &s3lib.HeadBucketInput{Bucket: aws.String(bucket)})
+	if err == nil {
+		return nil
+	}
+	var notFound *s3types.NotFound
+	if !errors.As(err, &notFound) {
+		return fmt.Errorf("head bucket %q: %w", bucket, err)
+	}
+	_, err = c.CreateBucket(context.Background(), &s3lib.CreateBucketInput{
+		Bucket:                    aws.String(bucket),
+		CreateBucketConfiguration: &s3types.CreateBucketConfiguration{},
+	})
+	if err != nil {
+		return fmt.Errorf("create bucket %q: %w", bucket, err)
+	}
+	fmt.Printf("cuobjtest: created bucket %q\n", bucket)
+	return nil
+}
+
+// printIterationResult renders one benchmark iteration according to mode.
+func printIterationResult(size int, r iterationResult) {
+	if *putOnly {
+		fmt.Printf("%-4d  %-14s %-10s\n", r.index, fmtDur(r.putDur), fmtGBps(size, r.putDur))
+		return
+	}
+	if *getOnly {
+		fmt.Printf("%-4d  %-14s %-10s\n", r.index, fmtDur(r.getDur), fmtGBps(size, r.getDur))
+		return
+	}
+	fmt.Printf("%-4d  %-14s %-10s   %-14s %-10s   %s\n",
+		r.index,
+		fmtDur(r.putDur), fmtGBps(size, r.putDur),
+		fmtDur(r.getDur), fmtGBps(size, r.getDur),
+		r.status)
+}
+
+// fmtDur renders a human-friendly duration.
+func fmtDur(d time.Duration) string {
+	ms := d.Seconds() * 1000
+	if ms >= 1000 {
+		return fmt.Sprintf("%.2f s", d.Seconds())
+	}
+	return fmt.Sprintf("%.2f ms", ms)
+}
+
+// fmtGBps formats bytes-per-second throughput as decimal GB/s.
+func fmtGBps(bytes int, d time.Duration) string {
+	if d <= 0 {
+		return "-"
+	}
+	return fmt.Sprintf("%.3f GB/s", float64(bytes)/d.Seconds()/1e9)
+}
+
+// formatSize renders a byte size using binary units when evenly divisible.
+func formatSize(n int) string {
+	switch {
+	case n >= 1<<30 && n%(1<<30) == 0:
+		return fmt.Sprintf("%d GiB", n>>30)
+	case n >= 1<<20 && n%(1<<20) == 0:
+		return fmt.Sprintf("%d MiB", n>>20)
+	case n >= 1<<10 && n%(1<<10) == 0:
+		return fmt.Sprintf("%d KiB", n>>10)
+	default:
+		return fmt.Sprintf("%d B", n)
+	}
+}
+
+// durStats returns average, minimum, and maximum durations.
+func durStats(ds []time.Duration) (avg, min, max time.Duration) {
+	min = time.Duration(math.MaxInt64)
+	var sum time.Duration
+	for _, d := range ds {
+		sum += d
+		if d < min {
+			min = d
+		}
+		if d > max {
+			max = d
+		}
+	}
+	avg = sum / time.Duration(len(ds))
+	return
+}
+
+// sumDurs returns the sum of the given durations.
+func sumDurs(ds []time.Duration) time.Duration {
+	var sum time.Duration
+	for _, d := range ds {
+		sum += d
+	}
+	return sum
+}
+
+// parseSize parses byte sizes like 64MiB, 1GiB, or 1000000.
+func parseSize(s string) (int, error) {
+	units := []struct {
+		suffix string
+		mult   int64
+	}{
+		{"GiB", 1 << 30}, {"GB", 1_000_000_000},
+		{"MiB", 1 << 20}, {"MB", 1_000_000},
+		{"KiB", 1 << 10}, {"KB", 1_000},
+		{"G", 1 << 30}, {"M", 1 << 20}, {"K", 1 << 10},
+		{"B", 1},
+	}
+	upper := strings.ToUpper(strings.TrimSpace(s))
+	for _, u := range units {
+		if strings.HasSuffix(upper, strings.ToUpper(u.suffix)) {
+			numStr := strings.TrimSpace(s[:len(s)-len(u.suffix)])
+			n, err := strconv.ParseFloat(numStr, 64)
+			if err != nil {
+				return 0, fmt.Errorf("invalid size %q", s)
+			}
+			return int(n * float64(u.mult)), nil
+		}
+	}
+	n, err := strconv.Atoi(strings.TrimSpace(s))
+	if err != nil {
+		return 0, fmt.Errorf("invalid size %q", s)
+	}
+	return n, nil
+}
+
+// fatalf prints an error and exits with status code 1.
+func fatalf(format string, args ...any) {
+	fmt.Fprintf(os.Stderr, format+"\n", args...)
+	os.Exit(1)
+}
@@ -0,0 +1,7 @@
+//go:build !cuobjclient_host
+
+package main
+
+func runModeLabel() string {
+	return "real-cuda-cuobject-token"
+}
@@ -0,0 +1,7 @@
+//go:build cuobjclient_host
+
+package main
+
+func runModeLabel() string {
+	return "host-memory-cuobject-token"
+}
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"bytes"
@@ -47,7 +47,9 @@ var (
 	allowInsecure bool
 )

-func adminCommand() *cli.Command {
+// AdminCommand returns the "admin" subcommand, common to all versitygw
+// binaries.
+func AdminCommand() *cli.Command {
 	return &cli.Command{
 		Name:        "admin",
 		Usage:       "admin CLI tool",
@@ -287,10 +289,10 @@ func getAdminCreds() (string, string, error) {

 	// Fallbacks to root user credentials
 	if access == "" {
-		access = rootUserAccess
+		access = RootUserAccess
 	}
 	if secret == "" {
-		secret = rootUserSecret
+		secret = RootUserSecret
 	}

 	if access == "" {
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"fmt"
@@ -25,7 +25,9 @@ var (
 	azAccount, azKey, azServiceURL, azSASToken string
 )

-func azureCommand() *cli.Command {
+// AzureCommand returns the "azure" subcommand, common to all versitygw
+// binaries.
+func AzureCommand() *cli.Command {
 	return &cli.Command{
 		Name:        "azure",
 		Usage:       "azure blob storage backend",
@@ -65,10 +67,10 @@ func azureCommand() *cli.Command {
 }

 func runAzure(ctx *cli.Context) error {
-	be, err := azure.New(azAccount, azKey, azServiceURL, azSASToken, copyObjectThreshold)
+	be, err := azure.New(azAccount, azKey, azServiceURL, azSASToken, CopyObjectThreshold)
 	if err != nil {
 		return fmt.Errorf("init azure: %w", err)
 	}

-	return runGateway(ctx.Context, be)
+	return RunGateway(ctx.Context, be)
 }
@@ -0,0 +1,46 @@
+// Copyright 2023 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// Package gwcli holds the versitygw CLI subcommands and state shared by all
+// versitygw-based main packages (cmd/versitygw, cmd/vgwrdma, ...). Hosting
+// binaries wire up their backend-specific flags/commands in their own main
+// package, and delegate the identical, backend-agnostic subcommands here.
+package gwcli
+
+import (
+	"context"
+
+	"github.com/versity/versitygw/backend"
+)
+
+// RootUserAccess and RootUserSecret are the root user credentials parsed by
+// the hosting binary's --access/--secret flags. The admin subcommand falls
+// back to these when no admin-specific credentials are provided.
+var (
+	RootUserAccess string
+	RootUserSecret string
+)
+
+// CopyObjectThreshold is the maximum allowed source object size in bytes for
+// CopyObject, populated by the hosting binary's --copy-object-threshold flag.
+var CopyObjectThreshold int64
+
+// DisableStrictBucketNames allows relaxed bucket naming, populated by the
+// hosting binary's --disable-strict-bucket-names flag.
+var DisableStrictBucketNames bool
+
+// RunGateway starts the S3 gateway server for the given backend. The hosting
+// binary's main package must set this before running any command that can
+// launch a backend (azure, plugin, s3).
+var RunGateway func(ctx context.Context, be backend.Backend) error
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"errors"
@@ -23,7 +23,9 @@ import (
 	"github.com/versity/versitygw/plugins"
 )

-func pluginCommand() *cli.Command {
+// PluginCommand returns the "plugin" subcommand, common to all versitygw
+// binaries.
+func PluginCommand() *cli.Command {
 	return &cli.Command{
 		Name:        "plugin",
 		Usage:       "load a backend from a plugin",
@@ -71,5 +73,5 @@ func runPluginBackend(ctx *cli.Context) error {
 		return err
 	}

-	return runGateway(ctx.Context, be)
+	return RunGateway(ctx.Context, be)
 }
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"fmt"
@@ -40,7 +40,9 @@ var (
 	dataIntegrityEtag    bool
 )

-func posixCommand() *cli.Command {
+// PosixCommand returns the "posix" subcommand, common to all versitygw
+// binaries.
+func PosixCommand() *cli.Command {
 	return &cli.Command{
 		Name:  "posix",
 		Usage: "posix filesystem storage backend",
@@ -176,10 +178,10 @@ func runPosix(ctx *cli.Context) error {
 		ForceNoTmpFile:       forceNoTmpFile,
 		ForceNoCopyFileRange: forceNoCopyFileRange,
 		EnableODirect:        enableODirect,
-		ValidateBucketNames:  disableStrictBucketNames,
+		ValidateBucketNames:  DisableStrictBucketNames,
 		Concurrency:          actionsConcurrency,
 		IOBufferSize:         ioBufferSize,
-		CopyObjectThreshold:  copyObjectThreshold,
+		CopyObjectThreshold:  CopyObjectThreshold,
 		DefaultEtag:          defaultEtag,
 		DataIntegrityEtag:    dataIntegrityEtag,
 	}
@@ -208,5 +210,5 @@ func runPosix(ctx *cli.Context) error {
 		return fmt.Errorf("failed to init posix backend: %w", err)
 	}

-	return runGateway(ctx.Context, be)
+	return RunGateway(ctx.Context, be)
 }
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"fmt"
@@ -36,7 +36,8 @@ var (
 	s3proxyGCSCompatibility          bool
 )

-func s3Command() *cli.Command {
+// S3Command returns the "s3" subcommand, common to all versitygw binaries.
+func S3Command() *cli.Command {
 	return &cli.Command{
 		Name:  "s3",
 		Usage: "s3 storage backend",
@@ -139,5 +140,5 @@ func runS3(ctx *cli.Context) error {
 	if err != nil {
 		return fmt.Errorf("init s3 backend: %w", err)
 	}
-	return runGateway(ctx.Context, be)
+	return RunGateway(ctx.Context, be)
 }
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"fmt"
@@ -29,7 +29,9 @@ var (
 	setProjectID     bool
 )

-func scoutfsCommand() *cli.Command {
+// ScoutfsCommand returns the "scoutfs" subcommand, common to all versitygw
+// binaries.
+func ScoutfsCommand() *cli.Command {
 	return &cli.Command{
 		Name:  "scoutfs",
 		Usage: "scoutfs filesystem storage backend",
@@ -143,10 +145,10 @@ func runScoutfs(ctx *cli.Context) error {
 	opts.NewDirPerm = fs.FileMode(dirPerms)
 	opts.DisableNoArchive = disableNoArchive
 	opts.VersioningDir = versioningDir
-	opts.ValidateBucketNames = disableStrictBucketNames
+	opts.ValidateBucketNames = DisableStrictBucketNames
 	opts.SetProjectID = setProjectID
 	opts.Concurrency = actionsConcurrency
-	opts.CopyObjectThreshold = copyObjectThreshold
+	opts.CopyObjectThreshold = CopyObjectThreshold
 	opts.DefaultEtag = defaultEtag
 	opts.DataIntegrityEtag = dataIntegrityEtag

@@ -155,5 +157,5 @@ func runScoutfs(ctx *cli.Context) error {
 		return fmt.Errorf("init scoutfs: %v", err)
 	}

-	return runGateway(ctx.Context, be)
+	return RunGateway(ctx.Context, be)
 }
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"fmt"
@@ -22,11 +22,15 @@ import (
 )

 var (
-	sigDone = make(chan struct{}, 1)
-	sigHup  = make(chan struct{}, 1)
+	// SigDone is signaled once on SIGINT/SIGTERM to begin shutdown.
+	SigDone = make(chan struct{}, 1)
+	// SigHup is signaled on every SIGHUP to trigger a config reload.
+	SigHup = make(chan struct{}, 1)
 )

-func setupSignalHandler() {
+// SetupSignalHandler starts a goroutine that translates SIGINT/SIGTERM into a
+// single SigDone notification and SIGHUP into repeated SigHup notifications.
+func SetupSignalHandler() {
 	sigs := make(chan os.Signal, 1)
 	signal.Notify(sigs, syscall.SIGINT, syscall.SIGTERM, syscall.SIGHUP)

@@ -35,9 +39,9 @@ func setupSignalHandler() {
 			fmt.Fprintf(os.Stderr, "caught signal %v\n", sig)
 			switch sig {
 			case syscall.SIGINT, syscall.SIGTERM:
-				sigDone <- struct{}{}
+				SigDone <- struct{}{}
 			case syscall.SIGHUP:
-				sigHup <- struct{}{}
+				SigHup <- struct{}{}
 			}
 		}
 	}()
@@ -12,7 +12,7 @@
 // specific language governing permissions and limitations
 // under the License.

-package main
+package gwcli

 import (
 	"encoding/json"
@@ -27,7 +27,9 @@ import (
 	"github.com/versity/versitygw/s3event"
 )

-func utilsCommand() *cli.Command {
+// UtilsCommand returns the "utils" subcommand, common to all versitygw
+// binaries.
+func UtilsCommand() *cli.Command {
 	return &cli.Command{
 		Name:  "utils",
 		Usage: "utility helper CLI tool",
@@ -11,6 +11,7 @@ import (

 	"github.com/versity/versitygw/backend/meta"
 	"github.com/versity/versitygw/backend/posix"
+	"github.com/versity/versitygw/cmd/internal/gwcli"
 	"github.com/versity/versitygw/tests/integration"
 )

@@ -28,14 +29,14 @@ func initEnv(dir string) {
 	region = "us-east-1"

 	// server
-	rootUserAccess = "user"
-	rootUserSecret = "pass"
+	gwcli.RootUserAccess = "user"
+	gwcli.RootUserSecret = "pass"
 	iamDir = dir
 	maxConnections = 250000
 	maxRequests = 100000
 	ports = []string{"127.0.0.1:7070"}
 	mpMaxParts = 10000
-	copyObjectThreshold = 5 * 1024 * 1024 * 1024
+	gwcli.CopyObjectThreshold = 5 * 1024 * 1024 * 1024

 	// client
 	awsID = "user"
@@ -24,6 +24,7 @@ import (

 	"github.com/urfave/cli/v2"
 	"github.com/versity/versitygw/backend"
+	"github.com/versity/versitygw/cmd/internal/gwcli"
 	"github.com/versity/versitygw/embedgw"
 	"github.com/versity/versitygw/s3api/utils"
 )
@@ -31,8 +32,6 @@ import (
 var (
 	ports                                  []string
 	admPorts                               []string
-	rootUserAccess                         string
-	rootUserSecret                         string
 	region                                 string
 	maxConnections, maxRequests            int
 	adminMaxConnections, adminMaxRequests  int
@@ -54,7 +53,6 @@ var (
 	pprof                                  string
 	quiet                                  bool
 	readonly                               bool
-	disableStrictBucketNames               bool
 	iamDir                                 string
 	ldapURL, ldapBindDN, ldapPassword      string
 	ldapQueryBase, ldapObjClasses          string
@@ -97,7 +95,6 @@ var (
 	websiteNoTLS                           bool
 	disableACLs                            bool
 	mpMaxParts                             int
-	copyObjectThreshold                    int64
 	socketPerm                             string
 )

@@ -111,24 +108,25 @@ var (
 )

 func main() {
-	setupSignalHandler()
+	gwcli.SetupSignalHandler()
+	gwcli.RunGateway = runGateway

 	app := initApp()

 	app.Commands = []*cli.Command{
-		posixCommand(),
-		scoutfsCommand(),
-		s3Command(),
-		azureCommand(),
-		pluginCommand(),
-		adminCommand(),
+		gwcli.PosixCommand(),
+		gwcli.ScoutfsCommand(),
+		gwcli.S3Command(),
+		gwcli.AzureCommand(),
+		gwcli.PluginCommand(),
+		gwcli.AdminCommand(),
 		testCommand(),
-		utilsCommand(),
+		gwcli.UtilsCommand(),
 	}

 	ctx, cancel := context.WithCancel(context.Background())
 	go func() {
-		<-sigDone
+		<-gwcli.SigDone
 		fmt.Fprintf(os.Stderr, "terminating signal caught, shutting down\n")
 		cancel()
 	}()
@@ -282,14 +280,14 @@ func initFlags() []cli.Flag {
 			Usage:       "root user access key",
 			EnvVars:     []string{"ROOT_ACCESS_KEY_ID", "ROOT_ACCESS_KEY"},
 			Aliases:     []string{"a"},
-			Destination: &rootUserAccess,
+			Destination: &gwcli.RootUserAccess,
 		},
 		&cli.StringFlag{
 			Name:        "secret",
 			Usage:       "root user secret access key",
 			EnvVars:     []string{"ROOT_SECRET_ACCESS_KEY", "ROOT_SECRET_KEY"},
 			Aliases:     []string{"s"},
-			Destination: &rootUserSecret,
+			Destination: &gwcli.RootUserSecret,
 		},
 		&cli.StringFlag{
 			Name:        "region",
@@ -732,7 +730,7 @@ func initFlags() []cli.Flag {
 			Name:        "disable-strict-bucket-names",
 			Usage:       "allow relaxed bucket naming (disables strict validation checks)",
 			EnvVars:     []string{"VGW_DISABLE_STRICT_BUCKET_NAMES"},
-			Destination: &disableStrictBucketNames,
+			Destination: &gwcli.DisableStrictBucketNames,
 		},
 		&cli.StringFlag{
 			Name:        "metrics-service-name",
@@ -797,7 +795,7 @@ func initFlags() []cli.Flag {
 			Usage:       "maximum allowed source object size in bytes for CopyObject; objects larger than this are rejected",
 			EnvVars:     []string{"VGW_COPY_OBJECT_THRESHOLD"},
 			Value:       5 * 1024 * 1024 * 1024,
-			Destination: &copyObjectThreshold,
+			Destination: &gwcli.CopyObjectThreshold,
 		},
 		&cli.StringFlag{
 			Name:        "socket-perm",
@@ -820,13 +818,13 @@ func runGateway(ctx context.Context, be backend.Backend) error {
 		}()
 	}

-	if copyObjectThreshold < 1 {
+	if gwcli.CopyObjectThreshold < 1 {
 		return fmt.Errorf("copy-object-threshold must be positive")
 	}

 	return embedgw.RunVersityGW(ctx, be, &embedgw.Config{
-		RootUserAccess:              rootUserAccess,
-		RootUserSecret:              rootUserSecret,
+		RootUserAccess:              gwcli.RootUserAccess,
+		RootUserSecret:              gwcli.RootUserSecret,
 		Region:                      region,
 		Ports:                       ports,
 		AdminPorts:                  admPorts,
@@ -846,7 +844,7 @@ func runGateway(ctx context.Context, be backend.Backend) error {
 		Readonly:                    readonly,
 		KeepAlive:                   keepAlive,
 		DisableACLs:                 disableACLs,
-		DisableStrictBucketNames:    disableStrictBucketNames,
+		DisableStrictBucketNames:    gwcli.DisableStrictBucketNames,
 		VirtualDomain:               virtualDomain,
 		HealthPath:                  healthPath,
 		SocketPerm:                  socketPerm,
@@ -919,7 +917,7 @@ func runGateway(ctx context.Context, be backend.Backend) error {
 		WebsiteCertFile:             websiteCertFile,
 		WebsiteKeyFile:              websiteKeyFile,
 		WebsiteNoTLS:                websiteNoTLS,
-		SigHup:                      sigHup,
+		SigHup:                      gwcli.SigHup,
 		Version:                     Version,
 		Build:                       Build,
 		BuildTime:                   BuildTime,
@@ -0,0 +1,359 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build linux && amd64 && cgo
+
+package cubackend
+
+// Package backend implements the cuObject-accelerated storage backend.
+// CuServer embeds the versitygw backend and overrides PutObject and
+// GetObject to use RDMA transfers when the request contains a cuObject
+// RDMA descriptor. All other S3 operations delegate directly to backend.
+
+import (
+	"bytes"
+	"context"
+	"fmt"
+	"io"
+	"net/http"
+
+	"github.com/aws/aws-sdk-go-v2/service/s3"
+
+	"github.com/versity/versitygw/backend"
+	"github.com/versity/versitygw/cumiddleware"
+	"github.com/versity/versitygw/debuglogger"
+	"github.com/versity/versitygw/rdma"
+	"github.com/versity/versitygw/rdma/bufferpool"
+	"github.com/versity/versitygw/s3err"
+	"github.com/versity/versitygw/s3response"
+)
+
+// CuServer is a versitygw Backend that embeds another backend and overrides
+// PutObject/GetObject for RDMA-accelerated transfers via cuObjServer.
+// All other S3 operations (list, head, delete, multipart, etc.) pass
+// through to the embedded backend unchanged.
+type CuServer struct {
+	backend.Backend
+	rdmaSrv *rdma.Server
+	pool    *bufferpool.Pool
+}
+
+// New creates a CuServer backend. It creates a cuObjServer instance,
+// starts the RDMA session, and pre-allocates a pool of RDMA-registered
+// host buffers.
+func New(opts CuServerOpts, be backend.Backend) (*CuServer, error) {
+	rdmaSrv, err := rdma.NewServer(opts.RDMAIP, opts.RDMAPort, opts.RDMATunables)
+	if err != nil {
+		be.Shutdown()
+		return nil, fmt.Errorf("cuserver: rdma server: %w", err)
+	}
+
+	// StartSession is a no-op when the library manages session start
+	// internally (e.g. libcuobjserver v1.2.0 calls startRDMASession from
+	// the cuObjServer constructor). It is kept here for forward compatibility
+	// with library versions that require an explicit call.
+	if err := rdmaSrv.StartSession(); err != nil {
+		rdmaSrv.Close()
+		be.Shutdown()
+		return nil, fmt.Errorf("cuserver: rdma session: %w", err)
+	}
+
+	bufSize := opts.PoolBufSize
+	if bufSize <= 0 {
+		bufSize = rdma.MaxTransferSize
+	}
+	bufCount := opts.PoolBufCount
+	if bufCount <= 0 {
+		bufCount = 4
+	}
+
+	pool, err := bufferpool.NewPool(rdmaSrv, bufSize, bufCount)
+	if err != nil {
+		rdmaSrv.Close()
+		be.Shutdown()
+		return nil, fmt.Errorf("cuserver: buffer pool: %w", err)
+	}
+
+	return &CuServer{
+		Backend: be,
+		rdmaSrv: rdmaSrv,
+		pool:    pool,
+	}, nil
+}
+
+// String returns a human-readable identifier for the backend.
+func (*CuServer) String() string {
+	return "cuObject Server"
+}
+
+// Shutdown closes the buffer pool and RDMA server, then shuts down the
+// embedded backend.
+func (cs *CuServer) Shutdown() {
+	cs.pool.Close()
+	cs.rdmaSrv.Close()
+	cs.Backend.Shutdown()
+}
+
+// PutObject handles S3 PUT requests. If the request context contains a
+// cuObject RDMA descriptor, data is transferred via RDMA READ from the
+// client's GPU memory into a local buffer, then written to the backend
+// Without an RDMA descriptor, delegates directly to backend.
+func (cs *CuServer) PutObject(ctx context.Context, po s3response.PutObjectInput) (s3response.PutObjectOutput, error) {
+	descr, ok := cumiddleware.GetRDMADescriptor(ctx)
+	if !ok {
+		debuglogger.Logf("doing normal put")
+		// Normal S3 path — no RDMA descriptor present
+		return cs.Backend.PutObject(ctx, po)
+	}
+
+	debuglogger.Logf("doing RDMA put")
+	// RDMA-accelerated path
+	size, ok := cumiddleware.GetRDMASize(ctx)
+	if !ok || size <= 0 {
+		// Size comes from the legacy 3-header scheme or, for the combined
+		// token scheme, the standard Content-Length header — neither was usable.
+		return s3response.PutObjectOutput{}, fmt.Errorf("cuserver: RDMA PUT requires a positive size via the %s header or a standard Content-Length header", cumiddleware.HeaderRDMASize)
+	}
+	if size > int64(rdma.MaxTransferSize) {
+		return s3response.PutObjectOutput{}, fmt.Errorf("cuserver: object size %d exceeds RDMA max %d", size, rdma.MaxTransferSize)
+	}
+
+	remoteStart := cumiddleware.GetRDMARemoteStart(ctx)
+	debuglogger.Logf("RDMA PUT params: key=%q size=%d remoteStart=0x%x descr=%s", keyFromPtr(po.Key), size, remoteStart, descr)
+
+	channelID, err := cs.rdmaSrv.AllocateChannel()
+	if err != nil {
+		return s3response.PutObjectOutput{}, fmt.Errorf("cuserver: alloc channel: %w", err)
+	}
+	defer cs.rdmaSrv.FreeChannel(channelID)
+
+	buf, _, err := cs.pool.Acquire(ctx)
+	if err != nil {
+		return s3response.PutObjectOutput{}, fmt.Errorf("cuserver: acquire buffer: %w", err)
+	}
+	defer cs.pool.Release(buf)
+
+	key := keyFromPtr(po.Key)
+
+	// Stream RDMA chunks into a pipe so backend can begin writing before
+	// the full transfer is complete.
+	pr, pw := io.Pipe()
+	po.Body = pr
+	po.ContentLength = &size
+	producerErrCh := make(chan error, 1)
+	go func() {
+		defer close(producerErrCh)
+
+		local := buf.Slice()
+		remaining := size
+		offset := int64(0)
+
+		for remaining > 0 {
+			chunk := rdmaPutChunkSize(len(local), remaining)
+			if chunk <= 0 {
+				wrapped := fmt.Errorf("cuserver: invalid RDMA PUT buffer size %d", len(local))
+				_ = pw.CloseWithError(wrapped)
+				producerErrCh <- wrapped
+				return
+			}
+
+			// RDMA READ current chunk from client into the start of local buffer.
+			n, e := cs.rdmaSrv.HandlePut(key, buf, remoteStart+uint64(offset), chunk, descr, channelID)
+			if e == nil && n != chunk {
+				e = fmt.Errorf("short RDMA transfer: got %d bytes, want %d", n, chunk)
+			}
+			if e != nil {
+				wrapped := fmt.Errorf("cuserver: RDMA PUT %q chunk offset=%d size=%d: %w", key, offset, chunk, e)
+				_ = pw.CloseWithError(wrapped)
+				producerErrCh <- wrapped
+				return
+			}
+
+			if _, e := pw.Write(local[:int(chunk)]); e != nil {
+				producerErrCh <- e
+				return
+			}
+
+			offset += chunk
+			remaining -= chunk
+		}
+
+		if e := pw.Close(); e != nil {
+			producerErrCh <- e
+			return
+		}
+		producerErrCh <- nil
+	}()
+
+	out, err := cs.Backend.PutObject(ctx, po)
+	if err != nil {
+		_ = pr.Close()
+		if prodErr := <-producerErrCh; prodErr != nil {
+			return s3response.PutObjectOutput{}, fmt.Errorf("cuserver: PUT %q failed (backend: %v; rdma/pipe: %w)", key, err, prodErr)
+		}
+		return s3response.PutObjectOutput{}, err
+	}
+
+	if prodErr := <-producerErrCh; prodErr != nil {
+		return s3response.PutObjectOutput{}, prodErr
+	}
+
+	cumiddleware.SetRDMAReplyHeader(ctx, http.StatusOK, size)
+	return out, nil
+}
+
+// UploadPart handles S3 multipart upload part requests. RDMA offload is not
+// supported here: object data would travel out-of-band via RDMA, but the
+// embedded backend's UploadPart still expects the part in the HTTP body.
+// Decline the descriptor so the client retries over the normal HTTP path
+// instead of silently uploading a wrong/empty part.
+func (cs *CuServer) UploadPart(ctx context.Context, input *s3.UploadPartInput) (*s3.UploadPartOutput, error) {
+	if _, ok := cumiddleware.GetRDMADescriptor(ctx); ok {
+		return nil, s3err.GetAPIError(s3err.ErrNotImplemented)
+	}
+	return cs.Backend.UploadPart(ctx, input)
+}
+
+// GetObject handles S3 GET requests. If the request context contains a
+// cuObject RDMA descriptor, the object is read from the backend
+// into an RDMA buffer, then transferred via RDMA WRITE to the client's
+// GPU memory. The HTTP response contains metadata only (no body).
+// Without an RDMA descriptor, delegates directly to backend.
+func (cs *CuServer) GetObject(ctx context.Context, input *s3.GetObjectInput) (*s3.GetObjectOutput, error) {
+	descr, ok := cumiddleware.GetRDMADescriptor(ctx)
+	if !ok {
+		debuglogger.Logf("doing normal get")
+		// Normal S3 path
+		return cs.Backend.GetObject(ctx, input)
+	}
+
+	debuglogger.Logf("doing RDMA get")
+
+	// First, get the object from backend (opens the file, returns metadata + Body)
+	res, err := cs.Backend.GetObject(ctx, input)
+	if err != nil {
+		return nil, err
+	}
+
+	// Determine the actual content size
+	var size int64
+	if res.ContentLength != nil {
+		size = *res.ContentLength
+	}
+	if size <= 0 {
+		// No content to transfer — return metadata only
+		return res, nil
+	}
+	if size > int64(rdma.MaxTransferSize) {
+		if res.Body != nil {
+			res.Body.Close()
+		}
+		return nil, fmt.Errorf("cuserver: object size %d exceeds RDMA max %d", size, rdma.MaxTransferSize)
+	}
+	if capSize, ok := cumiddleware.GetRDMASize(ctx); ok {
+		if capSize <= 0 {
+			if res.Body != nil {
+				res.Body.Close()
+			}
+			return nil, fmt.Errorf("cuserver: RDMA GET requires a positive client capacity")
+		}
+		if size > capSize {
+			if res.Body != nil {
+				res.Body.Close()
+			}
+			return nil, fmt.Errorf("cuserver: object/range size %d exceeds RDMA client capacity %d", size, capSize)
+		}
+	}
+
+	remoteStart := cumiddleware.GetRDMARemoteStart(ctx)
+
+	channelID, err := cs.rdmaSrv.AllocateChannel()
+	if err != nil {
+		res.Body.Close()
+		return nil, fmt.Errorf("cuserver: alloc channel: %w", err)
+	}
+	defer cs.rdmaSrv.FreeChannel(channelID)
+
+	buf, slice, err := cs.pool.Acquire(ctx)
+	if err != nil {
+		res.Body.Close()
+		return nil, fmt.Errorf("cuserver: acquire buffer: %w", err)
+	}
+	defer cs.pool.Release(buf)
+	defer res.Body.Close()
+
+	if len(slice) == 0 {
+		return nil, fmt.Errorf("cuserver: invalid RDMA GET buffer size %d", len(slice))
+	}
+
+	key := keyFromPtr(input.Key)
+	debuglogger.Logf("RDMA GET params: key=%q size=%d remoteStart=0x%x descr=%s", key, size, remoteStart, descr)
+
+	remaining := size
+	offset := int64(0)
+	for remaining > 0 {
+		chunk := rdmaPutChunkSize(len(slice), remaining)
+		if chunk <= 0 {
+			return nil, fmt.Errorf("cuserver: invalid RDMA GET buffer size %d", len(slice))
+		}
+
+		if _, err = io.ReadFull(res.Body, slice[:int(chunk)]); err != nil {
+			return nil, fmt.Errorf("cuserver: read object %q chunk offset=%d size=%d into RDMA buffer: %w", key, offset, chunk, err)
+		}
+
+		n, err := cs.rdmaSrv.HandleGet(key, buf, remoteStart+uint64(offset), chunk, descr, channelID)
+		if err == nil && n != chunk {
+			err = fmt.Errorf("short RDMA transfer: got %d bytes, want %d", n, chunk)
+		}
+		if err != nil {
+			return nil, fmt.Errorf("cuserver: RDMA GET %q chunk offset=%d size=%d: %w", key, offset, chunk, err)
+		}
+
+		offset += chunk
+		remaining -= chunk
+	}
+
+	// Return metadata only — the data was already sent via RDMA.
+	// Set Body to empty and clear ContentLength so the HTTP response
+	// carries only headers.
+	res.Body = io.NopCloser(bytes.NewReader(nil))
+	zero := int64(0)
+	res.ContentLength = &zero
+
+	cumiddleware.SetRDMAReplyHeader(ctx, http.StatusOK, size)
+	return res, nil
+}
+
+func keyFromPtr(p *string) string {
+	if p == nil {
+		return ""
+	}
+	return *p
+}
+
+// Ensure CuServer satisfies the Backend interface at compile time.
+var _ backend.Backend = (*CuServer)(nil)
+
+func rdmaPutChunkSize(bufSize int, remaining int64) int64 {
+	if bufSize <= 0 || remaining <= 0 {
+		return 0
+	}
+
+	chunkSize := int64(bufSize)
+	if remaining < chunkSize {
+		return remaining
+	}
+
+	return chunkSize
+}
@@ -0,0 +1,36 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+package cubackend
+
+import (
+	"github.com/versity/versitygw/rdma"
+)
+
+// CuServerOpts configures the CuServer backend.
+type CuServerOpts struct {
+	// RDMAIP is the server IP address for the RDMA interface.
+	RDMAIP string
+	// RDMAPort is the server port for the RDMA interface.
+	RDMAPort uint16
+
+	// Pool configuration
+	PoolBufSize  int // Size of each RDMA buffer (default: 1 GiB)
+	PoolBufCount int // Number of RDMA buffers to pre-allocate (default: 4)
+
+	// RDMATunables configures low-level RDMA connection parameters applied
+	// before the session starts. If nil, cuObjServer library defaults are
+	// used. Use rdma.DefaultRDMATunables() as a starting point.
+	RDMATunables *rdma.RDMATunables
+}
@@ -0,0 +1,36 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build !(linux && amd64 && cgo)
+
+package cubackend
+
+// Package backend provides the cuObject-accelerated storage backend.
+// This file is a stub for platforms without RDMA support.
+
+import (
+	"fmt"
+
+	"github.com/versity/versitygw/backend"
+)
+
+// CuServer is a non-functional stub on platforms without RDMA support.
+type CuServer struct {
+	backend.BackendUnsupported
+}
+
+// New always returns an unsupported-platform error on this build.
+func New(opts CuServerOpts, be backend.Backend) (*CuServer, error) {
+	return nil, fmt.Errorf("cuserver: RDMA backend not supported on this platform")
+}
@@ -0,0 +1,245 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+package cumiddleware
+
+// Package cumiddleware provides Fiber middleware for cuObject RDMA descriptor
+// extraction. The middleware reads cuObject-specific HTTP headers and stores
+// them as fasthttp user-values so the backend can detect RDMA-accelerated
+// requests.
+//
+// Important: versitygw passes ctx.RequestCtx() (*fasthttp.RequestCtx) to the
+// backend, not the fiber.Ctx or any context.WithValue wrapper. Values must
+// therefore be stored via RequestCtx.SetUserValue (string key) so that
+// context.Context.Value(stringKey) retrieves them correctly.
+
+import (
+	"context"
+	"errors"
+	"fmt"
+	"strconv"
+	"strings"
+
+	"github.com/gofiber/fiber/v3"
+	"github.com/valyala/fasthttp"
+)
+
+// String keys used for fasthttp user-value storage.
+// Must be string constants so *fasthttp.RequestCtx.Value(key) finds them.
+const (
+	localKeyRDMADescr       = "cuobj.rdma.descr"
+	localKeyRDMASize        = "cuobj.rdma.size"
+	localKeyRDMARemoteStart = "cuobj.rdma.remote_start"
+)
+
+// Header names used by the cuObject client to pass RDMA descriptor info.
+//
+// HeaderRDMADescr/HeaderRDMASize/HeaderRDMARemoteAddr are a legacy 3-header
+// scheme used only by this repo's own test tools (cmd/cuobjtest,
+// cmd/rdmatest). The real cuObject/minio-cpp SDK does not send these; it
+// sends a single combined HeaderRDMAToken instead. Both schemes are
+// supported so existing test tooling keeps working.
+const (
+	HeaderRDMADescr      = "X-CuObj-RDMA-Descr"
+	HeaderRDMASize       = "X-CuObj-Content-Length"
+	HeaderRDMARemoteAddr = "X-CuObj-Remote-Buf-Start"
+
+	// HeaderRDMAToken is the single combined header sent by the real
+	// cuObject client SDK (e.g. minio-cpp). Its value is a colon-delimited
+	// token per the cuObj RDMA descriptor protocol. The whole token is
+	// passed through verbatim to the RDMA backend as the descriptor.
+	//
+	// This is the canonical definition of the wire format; the encoder in
+	// cuwrapper/rdma_host_client_wrapper.cpp (build_token) must stay in
+	// sync with it. Fields, in order, colon-delimited, lowercase hex:
+	//
+	//   # | Field                                  | Type     | Width
+	//   --|-----------------------------------------|----------|-----------
+	//   1 | Remote base address (GPUMEM/SYSMEM)      | uint64   | 16 chars
+	//   2 | Max size of buffer region from base addr | uint32   | 8 chars
+	//   3 | Remote key (rkey)                        | uint32   | 8 chars
+	//   4 | LID of the client NIC                    | uint16   | 4 chars
+	//   5 | DCTN                                     | uint32   | 6 chars
+	//   6 | GID present (1|0)                        | bool     | 1 char
+	//   7 | GID of client NIC                        | 16 bytes | 32 chars
+	//
+	// Example: "0102030405060708:01020304:01020304:0102:010203:1:0102030405060708090a0b0c0d0e0f10"
+	HeaderRDMAToken = "X-Amz-Rdma-Token"
+
+	// HeaderRDMAReply is the response header sent after a successful
+	// RDMA-offloaded PUT/GET, per NVIDIA's documented cuObject workflow:
+	// "If the transfer is successfully offloaded to RDMA, the proxy responds
+	// with x-amz-rdma-reply." This implementation carries a numeric RDMA
+	// status code (e.g. HTTP-style 200/204/206 success classes) and is set
+	// only after the RDMA operation has actually succeeded.
+	HeaderRDMAReply = "X-Amz-Rdma-Reply"
+
+	// HeaderRDMABytesTransferred is the statistics header documented
+	// alongside HeaderRDMAToken/HeaderRDMAReply ("x-amz-rdma-bytes-transferred
+	// (for statistics)"). This header carries the numeric transferred-byte
+	// count and is set together with HeaderRDMAReply — see SetRDMAReplyHeader.
+	HeaderRDMABytesTransferred = "X-Amz-Rdma-Bytes-Transferred"
+)
+
+// CuObjMiddleware extracts cuObject RDMA headers and stores them in
+// the fasthttp request context so the backend can retrieve them via the
+// GetRDMA* helper functions.
+//
+// If neither the legacy descriptor header nor the combined RDMA token
+// header is present, the request passes through unchanged — the backend
+// will use the normal (non-RDMA) code path.
+//
+// If a descriptor is present but required fields are malformed, a 400 Bad
+// Request is returned immediately.
+func CuObjMiddleware(ctx fiber.Ctx) error {
+	descr := ctx.Get(HeaderRDMADescr)
+	token := ctx.Get(HeaderRDMAToken)
+	if descr == "" && token == "" {
+		return ctx.Next()
+	}
+
+	// Store directly on the underlying fasthttp RequestCtx so values survive
+	// the ctx.RequestCtx() call the versitygw controller uses to invoke the backend.
+	rctx := ctx.RequestCtx()
+
+	if descr != "" {
+		// Legacy 3-header scheme.
+		rctx.SetUserValue(localKeyRDMADescr, descr)
+
+		if sizeStr := ctx.Get(HeaderRDMASize); sizeStr != "" {
+			size, err := strconv.ParseInt(sizeStr, 10, 64)
+			if err != nil || size <= 0 {
+				return fiber.NewError(fiber.StatusBadRequest,
+					HeaderRDMASize+": must be a positive integer")
+			}
+			rctx.SetUserValue(localKeyRDMASize, size)
+		}
+
+		if addrStr := ctx.Get(HeaderRDMARemoteAddr); addrStr != "" {
+			addr, err := strconv.ParseUint(addrStr, 10, 64)
+			if err != nil {
+				return fiber.NewError(fiber.StatusBadRequest,
+					HeaderRDMARemoteAddr+": must be a non-negative integer")
+			}
+			rctx.SetUserValue(localKeyRDMARemoteStart, addr)
+		}
+
+		return ctx.Next()
+	}
+
+	// Combined token scheme (real cuObject client SDK). The descriptor
+	// passed to the RDMA backend is the raw token string; the remote base
+	// address is parsed from the token's first field, and the transfer size
+	// normally comes from the standard Content-Length header (RDMA replaces
+	// only the HTTP body, not the usual Content-Length semantics). A genuine
+	// RDMA control request can legitimately carry no HTTP body at all, in
+	// which case Content-Length is 0/absent; fall back to the token's own
+	// registered-buffer-size field (already part of the documented wire
+	// format) rather than leaving the backend with no usable size.
+	rctx.SetUserValue(localKeyRDMADescr, token)
+
+	remoteStart, err := parseRDMATokenBaseAddr(token)
+	if err != nil {
+		return fiber.NewError(fiber.StatusBadRequest, HeaderRDMAToken+": "+err.Error())
+	}
+	rctx.SetUserValue(localKeyRDMARemoteStart, remoteStart)
+
+	if size := rctx.Request.Header.ContentLength(); size > 0 {
+		rctx.SetUserValue(localKeyRDMASize, int64(size))
+	} else if bufSize, err := parseRDMATokenBufferSize(token); err == nil && bufSize > 0 {
+		rctx.SetUserValue(localKeyRDMASize, int64(bufSize))
+	}
+
+	return ctx.Next()
+}
+
+// parseRDMATokenBaseAddr extracts the remote base address — the first
+// colon-delimited field, a hex-encoded uint64 — from a cuObj RDMA token.
+func parseRDMATokenBaseAddr(token string) (uint64, error) {
+	i := 0
+	for i < len(token) && token[i] != ':' {
+		i++
+	}
+	if i == 0 || i == len(token) {
+		return 0, errors.New("malformed RDMA token: missing base address field")
+	}
+	addr, err := strconv.ParseUint(token[:i], 16, 64)
+	if err != nil {
+		return 0, fmt.Errorf("malformed RDMA token base address: %w", err)
+	}
+	return addr, nil
+}
+
+// parseRDMATokenBufferSize extracts the registered buffer size — the second
+// colon-delimited field, a hex-encoded uint32 — from a cuObj RDMA token. Used
+// as a size fallback when the request has no positive Content-Length.
+func parseRDMATokenBufferSize(token string) (uint32, error) {
+	fields := strings.SplitN(token, ":", 3)
+	if len(fields) < 2 || fields[1] == "" {
+		return 0, errors.New("malformed RDMA token: missing buffer size field")
+	}
+	size, err := strconv.ParseUint(fields[1], 16, 32)
+	if err != nil {
+		return 0, fmt.Errorf("malformed RDMA token buffer size: %w", err)
+	}
+	return uint32(size), nil
+}
+
+// GetRDMADescriptor retrieves the RDMA descriptor from the context.
+// Returns ("", false) if the request is not an RDMA-accelerated request.
+func GetRDMADescriptor(ctx context.Context) (string, bool) {
+	v, ok := ctx.Value(localKeyRDMADescr).(string)
+	return v, ok && v != ""
+}
+
+// GetRDMASize retrieves the RDMA content length from the context.
+func GetRDMASize(ctx context.Context) (int64, bool) {
+	v, ok := ctx.Value(localKeyRDMASize).(int64)
+	return v, ok
+}
+
+// GetRDMARemoteStart retrieves the remote buffer start address from the context.
+// Defaults to 0 if not set.
+func GetRDMARemoteStart(ctx context.Context) uint64 {
+	v, _ := ctx.Value(localKeyRDMARemoteStart).(uint64)
+	return v
+}
+
+// SetRDMAReplyHeader sets the HeaderRDMAReply (status code) and
+// HeaderRDMABytesTransferred (byte count) response headers, signaling to the
+// client that the transfer was completed via RDMA rather than the HTTP body.
+// Call only after the RDMA transfer has actually succeeded. ctx must be the
+// same *fasthttp.RequestCtx handed to the backend by versitygw; it is a no-op
+// otherwise (e.g. in unit tests without an HTTP layer).
+func SetRDMAReplyHeader(ctx context.Context, rdmaStatus int, transferredBytes int64) {
+	rctx, ok := ctx.(*fasthttp.RequestCtx)
+	if !ok {
+		return
+	}
+	rctx.Response.Header.Set(HeaderRDMAReply, strconv.Itoa(rdmaStatus))
+	rctx.Response.Header.Set(HeaderRDMABytesTransferred, strconv.FormatInt(transferredBytes, 10))
+}
+
+// InjectRDMAContext returns a copy of ctx with RDMA descriptor values set.
+// This bypasses the Fiber middleware and is intended for testing and direct
+// backend invocation without an HTTP layer.
+func InjectRDMAContext(ctx context.Context, descr string, size int64, remoteStart uint64) context.Context {
+	//lint:ignore SA1029 string keys required for fasthttp RequestCtx cross-package value lookup
+	ctx = context.WithValue(ctx, localKeyRDMADescr, descr)
+	//lint:ignore SA1029 string keys required for fasthttp RequestCtx cross-package value lookup
+	ctx = context.WithValue(ctx, localKeyRDMASize, size)
+	//lint:ignore SA1029 string keys required for fasthttp RequestCtx cross-package value lookup
+	ctx = context.WithValue(ctx, localKeyRDMARemoteStart, remoteStart)
+	return ctx
+}
@@ -0,0 +1,182 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+package cumiddleware
+
+// Protocol-level tests for the cuObject RDMA header parsing middleware.
+// These exercise the wire-format contract without requiring any RDMA
+// hardware: the middleware only ever inspects HTTP headers and stashes
+// parsed values on the request context.
+
+import (
+	"bytes"
+	"net/http"
+	"net/http/httptest"
+	"strconv"
+	"testing"
+
+	"github.com/gofiber/fiber/v3"
+	"github.com/stretchr/testify/assert"
+)
+
+// tokenWithBufSize constructs a valid cuObj RDMA token per the
+// HeaderRDMAToken wire format documented next to its definition.
+func tokenWithBufSize(baseAddr uint64, bufSize uint32) string {
+	return hex64(baseAddr) + ":" + hex32(bufSize) + ":01020304:0102:010203:1:0102030405060708090a0b0c0d0e0f10"
+}
+
+func hex64(v uint64) string {
+	s := strconv.FormatUint(v, 16)
+	for len(s) < 16 {
+		s = "0" + s
+	}
+	return s
+}
+
+func hex32(v uint32) string {
+	s := strconv.FormatUint(uint64(v), 16)
+	for len(s) < 8 {
+		s = "0" + s
+	}
+	return s
+}
+
+func newTestApp(t *testing.T) (*fiber.App, chan bool) {
+	t.Helper()
+	reached := make(chan bool, 1)
+	app := fiber.New()
+	app.Use("*", CuObjMiddleware)
+	app.Post("/", func(ctx fiber.Ctx) error {
+		reached <- true
+		return ctx.SendStatus(http.StatusOK)
+	})
+	return app, reached
+}
+
+func TestCuObjMiddlewareNoHeadersPassesThrough(t *testing.T) {
+	app, _ := newTestApp(t)
+	req := httptest.NewRequest(http.MethodPost, "/", nil)
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, http.StatusOK, resp.StatusCode)
+}
+
+func TestCuObjMiddlewareLegacyHeaders(t *testing.T) {
+	app := fiber.New()
+	app.Use("*", CuObjMiddleware)
+	app.Post("/", func(ctx fiber.Ctx) error {
+		rctx := ctx.RequestCtx()
+		descr, ok := GetRDMADescriptor(rctx)
+		assert.True(t, ok)
+		assert.Equal(t, "deadbeef", descr)
+		size, ok := GetRDMASize(rctx)
+		assert.True(t, ok)
+		assert.Equal(t, int64(4096), size)
+		assert.Equal(t, uint64(4660), GetRDMARemoteStart(rctx))
+		return ctx.SendStatus(http.StatusOK)
+	})
+
+	req := httptest.NewRequest(http.MethodPost, "/", nil)
+	req.Header.Set(HeaderRDMADescr, "deadbeef")
+	req.Header.Set(HeaderRDMASize, "4096")
+	req.Header.Set(HeaderRDMARemoteAddr, "4660")
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, http.StatusOK, resp.StatusCode)
+}
+
+func TestCuObjMiddlewareTokenWithContentLength(t *testing.T) {
+	app := fiber.New()
+	app.Use("*", CuObjMiddleware)
+	app.Post("/", func(ctx fiber.Ctx) error {
+		rctx := ctx.RequestCtx()
+		descr, ok := GetRDMADescriptor(rctx)
+		assert.True(t, ok)
+		token := tokenWithBufSize(0x1122334455667788, 0x2000)
+		assert.Equal(t, token, descr)
+		size, ok := GetRDMASize(rctx)
+		assert.True(t, ok)
+		assert.Equal(t, int64(1234), size) // from Content-Length, not the token buffer size
+		assert.Equal(t, uint64(0x1122334455667788), GetRDMARemoteStart(rctx))
+		return ctx.SendStatus(http.StatusOK)
+	})
+
+	token := tokenWithBufSize(0x1122334455667788, 0x2000)
+	body := make([]byte, 1234)
+	req := httptest.NewRequest(http.MethodPost, "/", bytes.NewReader(body))
+	req.Header.Set(HeaderRDMAToken, token)
+	req.ContentLength = int64(len(body))
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, http.StatusOK, resp.StatusCode)
+}
+
+// TestCuObjMiddlewareTokenZeroBodyFallsBackToTokenSize covers a genuine RDMA
+// control request that carries no HTTP body at all (Content-Length 0/unset):
+// the size must fall back to the token's own registered-buffer-size field
+// instead of leaving the backend with no usable size.
+func TestCuObjMiddlewareTokenZeroBodyFallsBackToTokenSize(t *testing.T) {
+	app := fiber.New()
+	app.Use("*", CuObjMiddleware)
+	app.Post("/", func(ctx fiber.Ctx) error {
+		rctx := ctx.RequestCtx()
+		size, ok := GetRDMASize(rctx)
+		assert.True(t, ok)
+		assert.Equal(t, int64(0x2000), size)
+		return ctx.SendStatus(http.StatusOK)
+	})
+
+	token := tokenWithBufSize(0x1122334455667788, 0x2000)
+	req := httptest.NewRequest(http.MethodPost, "/", nil)
+	req.Header.Set(HeaderRDMAToken, token)
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, http.StatusOK, resp.StatusCode)
+}
+
+func TestCuObjMiddlewareMalformedTokenRejected(t *testing.T) {
+	app, reached := newTestApp(t)
+	req := httptest.NewRequest(http.MethodPost, "/", nil)
+	req.Header.Set(HeaderRDMAToken, "not-a-valid-token")
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, http.StatusBadRequest, resp.StatusCode)
+	select {
+	case <-reached:
+		t.Fatal("handler should not have been reached for a malformed token")
+	default:
+	}
+}
+
+func TestSetRDMAReplyHeader(t *testing.T) {
+	app := fiber.New()
+	app.Post("/", func(ctx fiber.Ctx) error {
+		SetRDMAReplyHeader(ctx.RequestCtx(), http.StatusOK, 65536)
+		return ctx.SendStatus(http.StatusOK)
+	})
+
+	req := httptest.NewRequest(http.MethodPost, "/", nil)
+	resp, err := app.Test(req)
+	assert.NoError(t, err)
+	assert.Equal(t, "200", resp.Header.Get(HeaderRDMAReply))
+	assert.Equal(t, "65536", resp.Header.Get(HeaderRDMABytesTransferred))
+}
+
+func TestSetRDMAReplyHeaderNoopForNonFasthttpContext(t *testing.T) {
+	// Must not panic when called with a plain context.Context, e.g. from a
+	// unit test that injects RDMA values via InjectRDMAContext without an
+	// HTTP layer.
+	ctx := InjectRDMAContext(t.Context(), "descr", 10, 0)
+	SetRDMAReplyHeader(ctx, http.StatusOK, 10)
+}
@@ -0,0 +1,19 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+package cuobjclient
+
+// MaxTransferSize is the largest transfer size supported by the cuObject
+// client path in this package.
+const MaxTransferSize = 1 << 30
@@ -0,0 +1,163 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build linux && amd64 && cgo && cuobjclient_host
+
+// This file implements the cuObjClient Session API on top of the host-memory
+// RDMA client (package rdma/hostclient). It contains no CUDA/GPU dependency and
+// is selected with the `cuobjclient_host` build tag for RDMA-capable hosts that
+// have no GPU.
+//
+// The RDMA endpoint is configured from the environment so the NewSession(size)
+// SDK signature stays identical to the GPU build:
+//
+//	VGWRDMA_RDMA_DEV    RDMA device name (default: first device, e.g. mlx5_0)
+//	VGWRDMA_RDMA_PORT   HCA port number (default: 1)
+//	VGWRDMA_GID_INDEX   RoCE GID index from `ibv_devinfo -v`
+//	                    (default: auto-select first non-link-local GID)
+//	VGWRDMA_DC_KEY      Dynamic Connection key, decimal or 0x-hex
+//	                    (default: matches server DCKey 0xffeeddcc)
+package cuobjclient
+
+import (
+	"fmt"
+	"os"
+	"strconv"
+
+	s3lib "github.com/aws/aws-sdk-go-v2/service/s3"
+
+	"github.com/versity/versitygw/rdma/hostclient"
+)
+
+// Session owns a host buffer registered for cuObject RDMA transfers via the
+// host-memory RDMA client. Session methods are not safe for concurrent use.
+type Session struct {
+	client      *hostclient.Client
+	buf         []byte
+	token       string
+	size        int
+	remoteStart uint64
+}
+
+// NewSession creates a host-memory RDMA session for a fixed transfer size.
+func NewSession(size int) (*Session, error) {
+	if size <= 0 {
+		return nil, fmt.Errorf("invalid size %d", size)
+	}
+	if size > MaxTransferSize {
+		return nil, fmt.Errorf("invalid size %d: exceeds MaxTransferSize (%d)", size, MaxTransferSize)
+	}
+
+	dev := os.Getenv("VGWRDMA_RDMA_DEV")
+	port := envUint8("VGWRDMA_RDMA_PORT", 1)
+	gidIndex := envInt("VGWRDMA_GID_INDEX", -1)
+	dcKey := envUint64("VGWRDMA_DC_KEY", hostclient.DefaultDCKey)
+
+	client, err := hostclient.NewClient(dev, port, gidIndex, dcKey)
+	if err != nil {
+		return nil, err
+	}
+
+	buf, err := client.Register(size)
+	if err != nil {
+		client.Close()
+		return nil, err
+	}
+	token, err := client.Token()
+	if err != nil {
+		client.Close()
+		return nil, err
+	}
+
+	return &Session{
+		client:      client,
+		buf:         buf,
+		token:       token,
+		size:        size,
+		remoteStart: client.BufferAddr(),
+	}, nil
+}
+
+// Close releases the registered host buffer and closes the RDMA client.
+// After Close returns, the Session must not be used.
+func (s *Session) Close() {
+	if s.client == nil {
+		return
+	}
+	s.client.Close()
+	s.client = nil
+	s.buf = nil
+}
+
+// Upload copies src into the registered host buffer and performs a PUT. The
+// gateway RDMA-reads the buffer contents during the request.
+func (s *Session) Upload(base *s3lib.Client, bucket, key string, src []byte) error {
+	if len(src) != s.size {
+		return fmt.Errorf("upload size mismatch: got %d bytes, want %d", len(src), s.size)
+	}
+	copy(s.buf, src)
+	return doPut(base, bucket, key, int64(s.size), s.token, s.remoteStart)
+}
+
+// Download performs a GET; the gateway RDMA-writes into the registered host
+// buffer, then the bytes are copied into dst.
+func (s *Session) Download(base *s3lib.Client, bucket, key string, dst []byte) error {
+	if len(dst) != s.size {
+		return fmt.Errorf("download size mismatch: got %d bytes, want %d", len(dst), s.size)
+	}
+	for i := range s.buf {
+		s.buf[i] = 0
+	}
+	if err := doGet(base, bucket, key, int64(s.size), s.token, s.remoteStart); err != nil {
+		return err
+	}
+	copy(dst, s.buf)
+	return nil
+}
+
+func envUint8(name string, def uint8) uint8 {
+	v := os.Getenv(name)
+	if v == "" {
+		return def
+	}
+	n, err := strconv.ParseUint(v, 10, 8)
+	if err != nil {
+		return def
+	}
+	return uint8(n)
+}
+
+func envInt(name string, def int) int {
+	v := os.Getenv(name)
+	if v == "" {
+		return def
+	}
+	n, err := strconv.Atoi(v)
+	if err != nil {
+		return def
+	}
+	return n
+}
+
+func envUint64(name string, def uint64) uint64 {
+	v := os.Getenv(name)
+	if v == "" {
+		return def
+	}
+	n, err := strconv.ParseUint(v, 0, 64)
+	if err != nil {
+		return def
+	}
+	return n
+}
@@ -0,0 +1,210 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build linux && amd64 && cgo && !cuobjclient_host
+
+package cuobjclient
+
+/*
+#cgo CFLAGS: -I/usr/include -I${SRCDIR}/../cuwrapper
+#cgo LDFLAGS: -L${SRCDIR}/../rdma -l:libcuobjclientwrapper.a -L/usr/lib64 -lcuobjclient -lcudart -lstdc++ -ldl
+#include <stdlib.h>
+#include "cuobjclient_wrapper.h"
+*/
+import "C"
+
+import (
+	"fmt"
+	"unsafe"
+
+	s3lib "github.com/aws/aws-sdk-go-v2/service/s3"
+)
+
+const (
+	cuObjOpGet = 0
+	cuObjOpPut = 1
+)
+
+// Session owns a CUDA buffer registered for cuObject RDMA token exchange.
+type Session struct {
+	ctx         *C.cuobj_client_ctx_t
+	gpuBuf      unsafe.Pointer
+	size        int
+	remoteStart uint64
+}
+
+// NewSession creates a GPU-backed cuObject session for a fixed transfer size.
+// The returned session allocates CUDA memory, registers the descriptor with
+// cuobjclient, and prepares RDMA token generation for Upload and Download.
+// size must be in the range [1, MaxTransferSize].
+// A Session is reusable for multiple transfers as long as each Upload or
+// Download uses buffers with length equal to the size passed to NewSession.
+// Callers should keep a session open for repeated sequential operations and
+// call Close once the session is no longer needed.
+// Session methods are not safe for concurrent use. Do not call Upload,
+// Download, or Close from multiple goroutines at the same time on the same
+// Session value.
+// Call Close when finished to release all resources.
+func NewSession(size int) (*Session, error) {
+	if size <= 0 {
+		return nil, fmt.Errorf("invalid size %d", size)
+	}
+	if size > MaxTransferSize {
+		return nil, fmt.Errorf("invalid size %d: exceeds MaxTransferSize (%d)", size, MaxTransferSize)
+	}
+
+	ctx := C.cuobj_client_create()
+	if ctx == nil {
+		return nil, fmt.Errorf("cuobj_client_create failed")
+	}
+
+	s := &Session{ctx: ctx, size: size}
+	buf := C.cuobj_client_cuda_malloc(C.size_t(size))
+	if buf == nil {
+		s.Close()
+		return nil, fmt.Errorf("cudaMalloc(%d) failed", size)
+	}
+	s.gpuBuf = buf
+
+	if err := s.register(); err != nil {
+		s.Close()
+		return nil, err
+	}
+
+	s.remoteStart = uint64(C.cuobj_client_ptr_to_u64(s.gpuBuf))
+	return s, nil
+}
+
+// Close releases the GPU buffer and destroys the cuObject client context.
+// After Close returns, the Session must not be used.
+func (s *Session) Close() {
+	if s.ctx == nil {
+		return
+	}
+	if s.gpuBuf != nil {
+		_ = s.unregister()
+		_ = s.cudaFree()
+		s.gpuBuf = nil
+	}
+	C.cuobj_client_destroy(s.ctx)
+	s.ctx = nil
+}
+
+// Upload copies src into the session GPU buffer and performs a PUT operation
+// using cuObject RDMA headers.
+// src length must exactly match the size passed to NewSession.
+func (s *Session) Upload(base *s3lib.Client, bucket, key string, src []byte) error {
+	if len(src) != s.size {
+		return fmt.Errorf("upload size mismatch: got %d bytes, want %d", len(src), s.size)
+	}
+	if err := s.copyH2D(src); err != nil {
+		return err
+	}
+	token, err := s.getToken(cuObjOpPut)
+	if err != nil {
+		return err
+	}
+	defer s.putToken(token)
+	return doPut(base, bucket, key, int64(s.size), C.GoString(token), s.remoteStart)
+}
+
+// Download performs a GET operation into the session GPU buffer and copies the
+// resulting bytes back into dst.
+// dst length must exactly match the size passed to NewSession.
+func (s *Session) Download(base *s3lib.Client, bucket, key string, dst []byte) error {
+	if len(dst) != s.size {
+		return fmt.Errorf("download size mismatch: got %d bytes, want %d", len(dst), s.size)
+	}
+	if err := s.memset(0); err != nil {
+		return err
+	}
+	token, err := s.getToken(cuObjOpGet)
+	if err != nil {
+		return err
+	}
+	defer s.putToken(token)
+	if err := doGet(base, bucket, key, int64(s.size), C.GoString(token), s.remoteStart); err != nil {
+		return err
+	}
+	return s.copyD2H(dst)
+}
+
+func (s *Session) register() error {
+	rc := C.cuobj_client_register_descriptor(s.ctx, s.gpuBuf, C.size_t(s.size))
+	if rc != 0 {
+		return fmt.Errorf("cuMemObjGetDescriptor failed (rc=%d)", int(rc))
+	}
+	return nil
+}
+
+func (s *Session) unregister() error {
+	rc := C.cuobj_client_unregister_descriptor(s.ctx, s.gpuBuf)
+	if rc != 0 {
+		return fmt.Errorf("cuMemObjPutDescriptor failed (rc=%d)", int(rc))
+	}
+	return nil
+}
+
+func (s *Session) cudaFree() error {
+	rc := C.cuobj_client_cuda_free(s.gpuBuf)
+	if rc != 0 {
+		return fmt.Errorf("cudaFree failed: %s", C.GoString(C.cuobj_client_cuda_error_string(rc)))
+	}
+	return nil
+}
+
+func (s *Session) memset(val int) error {
+	rc := C.cuobj_client_cuda_memset(s.gpuBuf, C.int(val), C.size_t(s.size))
+	if rc != 0 {
+		return fmt.Errorf("cudaMemset failed: %s", C.GoString(C.cuobj_client_cuda_error_string(rc)))
+	}
+	return nil
+}
+
+func (s *Session) copyH2D(src []byte) error {
+	if len(src) == 0 {
+		return nil
+	}
+	rc := C.cuobj_client_cuda_memcpy_h2d(s.gpuBuf, unsafe.Pointer(&src[0]), C.size_t(len(src)))
+	if rc != 0 {
+		return fmt.Errorf("cudaMemcpy H2D failed: %s", C.GoString(C.cuobj_client_cuda_error_string(rc)))
+	}
+	return nil
+}
+
+func (s *Session) copyD2H(dst []byte) error {
+	if len(dst) == 0 {
+		return nil
+	}
+	rc := C.cuobj_client_cuda_memcpy_d2h(unsafe.Pointer(&dst[0]), s.gpuBuf, C.size_t(len(dst)))
+	if rc != 0 {
+		return fmt.Errorf("cudaMemcpy D2H failed: %s", C.GoString(C.cuobj_client_cuda_error_string(rc)))
+	}
+	return nil
+}
+
+func (s *Session) getToken(op int) (*C.char, error) {
+	t := C.cuobj_client_get_rdma_token(s.ctx, s.gpuBuf, C.size_t(s.size), C.size_t(0), C.int(op))
+	if t == nil {
+		return nil, fmt.Errorf("cuMemObjGetRDMAToken failed")
+	}
+	return t, nil
+}
+
+func (s *Session) putToken(token *C.char) {
+	if token == nil {
+		return
+	}
+	_ = C.cuobj_client_put_rdma_token(s.ctx, token)
+}
@@ -0,0 +1,66 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build !(linux && amd64 && cgo)
+
+package cuobjclient
+
+import (
+	"fmt"
+
+	s3lib "github.com/aws/aws-sdk-go-v2/service/s3"
+)
+
+// Session is a non-Linux stub so packages compile on unsupported platforms.
+type Session struct{}
+
+// NewSession always returns an unsupported-platform error on this build.
+//
+// On supported platforms, sessions are intended to be reused for multiple
+// sequential transfers and are not safe for concurrent use by multiple
+// goroutines.
+func NewSession(size int) (*Session, error) {
+	if size <= 0 {
+		return nil, fmt.Errorf("invalid size %d", size)
+	}
+	if size > MaxTransferSize {
+		return nil, fmt.Errorf("invalid size %d: exceeds MaxTransferSize (%d)", size, MaxTransferSize)
+	}
+	return nil, fmt.Errorf("cuobjclient: NewSession is only supported on linux/amd64 with cgo")
+}
+
+// Close is a no-op in the unsupported-platform stub.
+func (s *Session) Close() {
+	_ = s
+}
+
+// Upload always returns an unsupported-platform error on this build.
+func (s *Session) Upload(base *s3lib.Client, bucket, key string, src []byte) error {
+	_ = s
+	_ = base
+	_ = bucket
+	_ = key
+	_ = src
+	return fmt.Errorf("cuobjclient: Upload is only supported on linux/amd64 with cgo")
+}
+
+// Download always returns an unsupported-platform error on this build.
+func (s *Session) Download(base *s3lib.Client, bucket, key string, dst []byte) error {
+	_ = s
+	_ = base
+	_ = bucket
+	_ = key
+	_ = dst
+	return fmt.Errorf("cuobjclient: Download is only supported on linux/amd64 with cgo")
+}
@@ -0,0 +1,164 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build linux && amd64 && cgo
+
+// This file holds the S3 request plumbing shared by the GPU (session_linux.go)
+// and host-memory (session_host_linux.go) session implementations: both issue
+// a zero-byte PUT/GET carrying the cuObject RDMA descriptor headers, and let
+// the gateway perform the actual transfer via RDMA.
+package cuobjclient
+
+import (
+	"context"
+	"fmt"
+	"io"
+	"net/http"
+	"strconv"
+	"strings"
+
+	"github.com/aws/aws-sdk-go-v2/aws"
+	smithymiddleware "github.com/aws/smithy-go/middleware"
+	smithyhttp "github.com/aws/smithy-go/transport/http"
+
+	s3lib "github.com/aws/aws-sdk-go-v2/service/s3"
+
+	"github.com/versity/versitygw/cumiddleware"
+)
+
+func doPut(base *s3lib.Client, bucket, key string, size int64, descr string, remoteStart uint64) error {
+	var replyStatus string
+	var transferredHeader string
+	c := withRDMAHeaders(base, descr, size, remoteStart, &replyStatus, &transferredHeader)
+	_, err := c.PutObject(context.Background(), &s3lib.PutObjectInput{
+		Bucket:        aws.String(bucket),
+		Key:           aws.String(key),
+		Body:          strings.NewReader(""),
+		ContentLength: aws.Int64(0),
+	})
+	if err != nil {
+		return err
+	}
+	if replyStatus == "" {
+		return fmt.Errorf("cuobjclient: gateway did not confirm RDMA offload (missing %s response header)", cumiddleware.HeaderRDMAReply)
+	}
+	rdmaStatus, err := strconv.Atoi(replyStatus)
+	if err != nil {
+		return fmt.Errorf("cuobjclient: invalid %s header %q: %w", cumiddleware.HeaderRDMAReply, replyStatus, err)
+	}
+	if rdmaStatus != http.StatusOK && rdmaStatus != http.StatusNoContent {
+		return fmt.Errorf("cuobjclient: RDMA offload not successful, %s=%d", cumiddleware.HeaderRDMAReply, rdmaStatus)
+	}
+	if transferredHeader == "" {
+		return fmt.Errorf("cuobjclient: missing %s response header", cumiddleware.HeaderRDMABytesTransferred)
+	}
+	transferred, err := strconv.ParseInt(transferredHeader, 10, 64)
+	if err != nil {
+		return fmt.Errorf("cuobjclient: invalid %s header %q: %w", cumiddleware.HeaderRDMABytesTransferred, transferredHeader, err)
+	}
+	if transferred != size {
+		return fmt.Errorf("cuobjclient: RDMA transferred %d bytes, want %d", transferred, size)
+	}
+	return nil
+}
+
+func doGet(base *s3lib.Client, bucket, key string, size int64, descr string, remoteStart uint64) error {
+	var replyStatus string
+	var transferredHeader string
+	c := withRDMAHeaders(base, descr, size, remoteStart, &replyStatus, &transferredHeader)
+	out, err := c.GetObject(context.Background(), &s3lib.GetObjectInput{
+		Bucket: aws.String(bucket),
+		Key:    aws.String(key),
+	})
+	if err != nil {
+		return err
+	}
+	defer out.Body.Close()
+	// The RDMA GET path deliberately reports ContentLength 0 on the HTTP
+	// response (the object bytes were already sent via RDMA, not the HTTP
+	// body), so the transfer must be confirmed via the gateway's RDMA reply
+	// header instead. A short/oversized transfer or a missing reply (offload
+	// silently not applied) is caught here rather than treated as success.
+	if replyStatus == "" {
+		return fmt.Errorf("cuobjclient: gateway did not confirm RDMA offload (missing %s response header)", cumiddleware.HeaderRDMAReply)
+	}
+	rdmaStatus, err := strconv.Atoi(replyStatus)
+	if err != nil {
+		return fmt.Errorf("cuobjclient: invalid %s header %q: %w", cumiddleware.HeaderRDMAReply, replyStatus, err)
+	}
+	if rdmaStatus != http.StatusOK && rdmaStatus != http.StatusNoContent && rdmaStatus != http.StatusPartialContent {
+		return fmt.Errorf("cuobjclient: RDMA offload not successful, %s=%d", cumiddleware.HeaderRDMAReply, rdmaStatus)
+	}
+	if transferredHeader == "" {
+		return fmt.Errorf("cuobjclient: missing %s response header", cumiddleware.HeaderRDMABytesTransferred)
+	}
+	transferred, err := strconv.ParseInt(transferredHeader, 10, 64)
+	if err != nil {
+		return fmt.Errorf("cuobjclient: invalid %s header %q: %w", cumiddleware.HeaderRDMABytesTransferred, transferredHeader, err)
+	}
+	if transferred != size {
+		return fmt.Errorf("cuobjclient: RDMA transferred %d bytes, want %d", transferred, size)
+	}
+	_, err = io.Copy(io.Discard, out.Body)
+	return err
+}
+
+// withRDMAHeaders returns a client that adds the legacy RDMA descriptor,
+// size, and remote-address headers to every request. The headers are added
+// via a Build-step middleware — which runs before the Finalize step that
+// signs the request — so SigV4 covers them in SignedHeaders; an
+// intermediary can no longer alter the RDMA controls without invalidating
+// the signature. The caller's HTTPClient (with its own TLS/proxy/timeout
+// configuration) is left untouched. Automatic request/response checksum
+// calculation is disabled for these control requests: the SDK would
+// otherwise checksum the empty HTTP body instead of the actual RDMA payload.
+// If replyStatus/transferred are non-nil, they are set to the response
+// HeaderRDMAReply and HeaderRDMABytesTransferred values (empty if absent).
+func withRDMAHeaders(base *s3lib.Client, descr string, size int64, remoteStart uint64, replyStatus, transferred *string) *s3lib.Client {
+	opts := base.Options()
+	opts.RequestChecksumCalculation = aws.RequestChecksumCalculationWhenRequired
+	opts.ResponseChecksumValidation = aws.ResponseChecksumValidationWhenRequired
+	opts.APIOptions = append(opts.APIOptions, func(stack *smithymiddleware.Stack) error {
+		if err := stack.Build.Add(smithymiddleware.BuildMiddlewareFunc("AddRDMAHeaders",
+			func(ctx context.Context, in smithymiddleware.BuildInput, next smithymiddleware.BuildHandler) (
+				smithymiddleware.BuildOutput, smithymiddleware.Metadata, error) {
+				if req, ok := in.Request.(*smithyhttp.Request); ok {
+					req.Header.Set(cumiddleware.HeaderRDMADescr, descr)
+					req.Header.Set(cumiddleware.HeaderRDMASize, strconv.FormatInt(size, 10))
+					req.Header.Set(cumiddleware.HeaderRDMARemoteAddr, strconv.FormatUint(remoteStart, 10))
+				}
+				return next.HandleBuild(ctx, in)
+			}), smithymiddleware.Before); err != nil {
+			return err
+		}
+		if replyStatus == nil && transferred == nil {
+			return nil
+		}
+		return stack.Deserialize.Add(smithymiddleware.DeserializeMiddlewareFunc("CaptureRDMAReply",
+			func(ctx context.Context, in smithymiddleware.DeserializeInput, next smithymiddleware.DeserializeHandler) (
+				smithymiddleware.DeserializeOutput, smithymiddleware.Metadata, error) {
+				out, metadata, err := next.HandleDeserialize(ctx, in)
+				if resp, ok := out.RawResponse.(*smithyhttp.Response); ok {
+					if replyStatus != nil {
+						*replyStatus = resp.Header.Get(cumiddleware.HeaderRDMAReply)
+					}
+					if transferred != nil {
+						*transferred = resp.Header.Get(cumiddleware.HeaderRDMABytesTransferred)
+					}
+				}
+				return out, metadata, err
+			}), smithymiddleware.After)
+	})
+	return s3lib.New(opts)
+}
@@ -0,0 +1,153 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// C wrapper implementation for cuObjClient and CUDA runtime operations.
+
+#include "cuobjclient_wrapper.h"
+
+#include <cerrno>
+#include <cstddef>
+#include <cstdint>
+
+#include <cuda_runtime_api.h>
+#include <cuobjclient.h>
+
+struct cuobj_client_ctx {
+    CUObjOps_t ops;
+    cuObjClient *client;
+};
+
+static ssize_t noop_get(const void *, char *, size_t, loff_t, const cufileRDMAInfo_t *) {
+    return -EOPNOTSUPP;
+}
+
+static ssize_t noop_put(const void *, const char *, size_t, loff_t, const cufileRDMAInfo_t *) {
+    return -EOPNOTSUPP;
+}
+
+extern "C" {
+
+cuobj_client_ctx_t* cuobj_client_create(void) {
+    try {
+        auto *ctx = new cuobj_client_ctx_t();
+        ctx->ops.get = noop_get;
+        ctx->ops.put = noop_put;
+        ctx->client = new cuObjClient(ctx->ops, CUOBJ_PROTO_RDMA_DC_V1);
+        return ctx;
+    } catch (...) {
+        return nullptr;
+    }
+}
+
+void cuobj_client_destroy(cuobj_client_ctx_t *ctx) {
+    if (!ctx) {
+        return;
+    }
+    delete ctx->client;
+    delete ctx;
+}
+
+void* cuobj_client_cuda_malloc(size_t size) {
+    void *ptr = nullptr;
+    cudaError_t rc = cudaMalloc(&ptr, size);
+    if (rc != cudaSuccess) {
+        return nullptr;
+    }
+    return ptr;
+}
+
+int cuobj_client_cuda_free(void *ptr) {
+    cudaError_t rc = cudaFree(ptr);
+    return static_cast<int>(rc);
+}
+
+int cuobj_client_cuda_memset(void *ptr, int value, size_t size) {
+    cudaError_t rc = cudaMemset(ptr, value, size);
+    return static_cast<int>(rc);
+}
+
+int cuobj_client_cuda_memcpy_h2d(void *dst_dev, const void *src_host, size_t size) {
+    cudaError_t rc = cudaMemcpy(dst_dev, src_host, size, cudaMemcpyHostToDevice);
+    return static_cast<int>(rc);
+}
+
+int cuobj_client_cuda_memcpy_d2h(void *dst_host, const void *src_dev, size_t size) {
+    cudaError_t rc = cudaMemcpy(dst_host, src_dev, size, cudaMemcpyDeviceToHost);
+    return static_cast<int>(rc);
+}
+
+const char* cuobj_client_cuda_error_string(int cuda_err) {
+    return cudaGetErrorString(static_cast<cudaError_t>(cuda_err));
+}
+
+int cuobj_client_register_descriptor(cuobj_client_ctx_t *ctx, void *ptr, size_t size) {
+    if (!ctx || !ctx->client || !ptr || size == 0) {
+        return 1;
+    }
+    cuObjErr_t rc = ctx->client->cuMemObjGetDescriptor(ptr, size);
+    return static_cast<int>(rc);
+}
+
+int cuobj_client_unregister_descriptor(cuobj_client_ctx_t *ctx, void *ptr) {
+    if (!ctx || !ctx->client || !ptr) {
+        return 1;
+    }
+    cuObjErr_t rc = ctx->client->cuMemObjPutDescriptor(ptr);
+    // Synchronize the device to ensure the hardware-level RDMA deregistration
+    // (memory unpinning, hardware lock release) completes before this call
+    // returns. Without this, a subsequent process that immediately calls
+    // cuMemObjGetDescriptor on a new allocation can race with the driver's
+    // asynchronous cleanup and receive a hardware rejection.
+    cudaDeviceSynchronize();
+    return static_cast<int>(rc);
+}
+
+char* cuobj_client_get_rdma_token(cuobj_client_ctx_t *ctx,
+                                  void *ptr,
+                                  size_t size,
+                                  size_t buffer_offset,
+                                  int operation) {
+    if (!ctx || !ctx->client || !ptr || size == 0) {
+        return nullptr;
+    }
+    cuObjOpType_t op;
+    if (operation == CUOBJCLIENT_OP_GET) {
+        op = CUOBJ_GET;
+    } else if (operation == CUOBJCLIENT_OP_PUT) {
+        op = CUOBJ_PUT;
+    } else {
+        return nullptr;
+    }
+
+    char *token = nullptr;
+    cuObjErr_t rc = ctx->client->cuMemObjGetRDMAToken(ptr, size, buffer_offset, op, &token);
+    if (rc != CU_OBJ_SUCCESS) {
+        return nullptr;
+    }
+    return token;
+}
+
+int cuobj_client_put_rdma_token(cuobj_client_ctx_t *ctx, char *token) {
+    if (!ctx || !ctx->client || !token) {
+        return 1;
+    }
+    cuObjErr_t rc = ctx->client->cuMemObjPutRDMAToken(token);
+    return static_cast<int>(rc);
+}
+
+uint64_t cuobj_client_ptr_to_u64(void *ptr) {
+    return static_cast<uint64_t>(reinterpret_cast<uintptr_t>(ptr));
+}
+
+} // extern "C"
@@ -0,0 +1,67 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// C wrapper for cuObjClient + CUDA runtime APIs.
+// Exposes a C ABI suitable for CGO.
+
+#ifndef CUOBJCLIENT_WRAPPER_H
+#define CUOBJCLIENT_WRAPPER_H
+
+#include <stddef.h>
+#include <stdint.h>
+
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+typedef struct cuobj_client_ctx cuobj_client_ctx_t;
+
+// cuObj operation values (match cuObjOpType_t in cuobjclient headers).
+#define CUOBJCLIENT_OP_GET 0
+#define CUOBJCLIENT_OP_PUT 1
+
+// Lifecycle
+cuobj_client_ctx_t* cuobj_client_create(void);
+void                cuobj_client_destroy(cuobj_client_ctx_t *ctx);
+
+// GPU memory management
+void*       cuobj_client_cuda_malloc(size_t size);
+int         cuobj_client_cuda_free(void *ptr);
+int         cuobj_client_cuda_memset(void *ptr, int value, size_t size);
+int         cuobj_client_cuda_memcpy_h2d(void *dst_dev, const void *src_host, size_t size);
+int         cuobj_client_cuda_memcpy_d2h(void *dst_host, const void *src_dev, size_t size);
+const char* cuobj_client_cuda_error_string(int cuda_err);
+
+// cuObject memory registration
+int cuobj_client_register_descriptor(cuobj_client_ctx_t *ctx, void *ptr, size_t size);
+int cuobj_client_unregister_descriptor(cuobj_client_ctx_t *ctx, void *ptr);
+
+// RDMA token management.
+// Returns an allocated descriptor string on success, NULL on failure.
+// Caller must release it via cuobj_client_put_rdma_token.
+char* cuobj_client_get_rdma_token(cuobj_client_ctx_t *ctx,
+                                  void *ptr,
+                                  size_t size,
+                                  size_t buffer_offset,
+                                  int operation);
+int   cuobj_client_put_rdma_token(cuobj_client_ctx_t *ctx, char *token);
+
+// Utility
+uint64_t cuobj_client_ptr_to_u64(void *ptr);
+
+#ifdef __cplusplus
+}
+#endif
+
+#endif // CUOBJCLIENT_WRAPPER_H
@@ -0,0 +1,247 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// C wrapper implementation for cuObjServer.
+// Bridges extern "C" functions to the C++ cuObjServer class.
+
+#include "cuobjserver_wrapper.h"
+#include "cuobjserver.h"
+
+#include <dlfcn.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <atomic>
+#include <iostream>
+#include <string>
+
+// ---------------------------------------------------------------------------
+// Session management — runtime symbol resolution
+//
+// libcuobjserver.so may export startRDMASession / closeRDMASession under one
+// of two C++ mangled names depending on when the library was compiled:
+//
+//   Newer header layout (RDMAConnection base class):
+//     _ZN14RDMAConnection16startRDMASessionEv
+//     _ZN14RDMAConnection16closeRDMASessionEv
+//
+//   Older layout (method directly on cuObjServer):
+//     _ZN11cuObjServer16startRDMASessionEv
+//     _ZN11cuObjServer16closeRDMASessionEv
+//
+// We resolve lazily at first call so neither name needs to be present at
+// link time, and the code works with both library generations.
+// ---------------------------------------------------------------------------
+
+typedef int  (*rdma_start_fn_t)(void *);
+typedef void (*rdma_close_fn_t)(void *);
+
+// Verbose wrapper logs are enabled when cuobj_server_set_telem_flags includes
+// info/debug bits (configured by cuserver -debug).
+static std::atomic<bool> g_verbose_logs{false};
+
+static rdma_start_fn_t find_start_rdma_session() {
+    void *proc = dlopen(nullptr, RTLD_LAZY);
+    if (!proc) return nullptr;
+    rdma_start_fn_t fn = reinterpret_cast<rdma_start_fn_t>(
+        dlsym(proc, "_ZN14RDMAConnection16startRDMASessionEv"));
+    if (!fn)
+        fn = reinterpret_cast<rdma_start_fn_t>(
+            dlsym(proc, "_ZN11cuObjServer16startRDMASessionEv"));
+    dlclose(proc);
+    return fn;
+}
+
+static rdma_close_fn_t find_close_rdma_session() {
+    void *proc = dlopen(nullptr, RTLD_LAZY);
+    if (!proc) return nullptr;
+    rdma_close_fn_t fn = reinterpret_cast<rdma_close_fn_t>(
+        dlsym(proc, "_ZN14RDMAConnection16closeRDMASessionEv"));
+    if (!fn)
+        fn = reinterpret_cast<rdma_close_fn_t>(
+            dlsym(proc, "_ZN11cuObjServer16closeRDMASessionEv"));
+    dlclose(proc);
+    return fn;
+}
+
+extern "C" {
+
+cuobj_server_t* cuobj_server_create(const char *ip, unsigned short port, unsigned proto) {
+    try {
+        auto *srv = new cuObjServer(ip, port, proto);
+        return reinterpret_cast<cuobj_server_t*>(srv);
+    } catch (...) {
+        return nullptr;
+    }
+}
+
+// Build a cuObjRDMATunable from the flat C struct, shared by
+// cuobj_server_create_with_config and cuobj_server_init_rdma_config.
+static cuObjRDMATunable tunables_from_c(const cuobj_rdma_tunables_t *t) {
+    cuObjRDMATunable config;
+    config.setNumDcis(t->num_dcis);
+    config.setCqDepth(t->cq_depth);
+    config.setDcKey(t->dc_key);
+    config.setServiceLevel(t->service_level);
+    config.setTimeout(t->timeout);
+    config.setHopLimit(t->hop_limit);
+    config.setPkeyIndex(t->pkey_index);
+    config.setDelayInterval(t->delay_interval);
+    config.setDelayMode(static_cast<cuObjDelayMode_t>(t->delay_mode));
+    config.setRetryCount(t->retry_cnt);
+    config.setQPResetOnFailure(t->qp_reset_on_failure != 0);
+    config.setTrafficClass(t->traffic_class);
+    config.setMaxRdAtomic(t->max_rd_atomic);
+    return config;
+}
+
+cuobj_server_t* cuobj_server_create_with_config(const char *ip, unsigned short port,
+                                                unsigned proto,
+                                                const cuobj_rdma_tunables_t *t) {
+    try {
+        cuObjRDMATunable config = tunables_from_c(t);
+        auto *srv = new cuObjServer(ip, port, proto, config);
+        return reinterpret_cast<cuobj_server_t*>(srv);
+    } catch (...) {
+        return nullptr;
+    }
+}
+
+void cuobj_server_destroy(cuobj_server_t *srv) {
+    delete reinterpret_cast<cuObjServer*>(srv);
+}
+
+int cuobj_server_start_session(cuobj_server_t *srv) {
+    static rdma_start_fn_t fn = find_start_rdma_session();
+    if (!fn) {
+        // Symbol not exported — library calls startRDMASession() internally
+        // from the cuObjServer constructor. Verify the session actually came
+        // up instead of unconditionally reporting success.
+        auto *s = reinterpret_cast<cuObjServer*>(srv);
+        return s->isConnected() ? 0 : -1;
+    }
+    int rc = fn(reinterpret_cast<void *>(srv));
+    if (rc != 0)
+        fprintf(stderr, "cuobjwrapper: startRDMASession returned %d\n", rc);
+    return rc;
+}
+
+void cuobj_server_close_session(cuobj_server_t *srv) {
+    static rdma_close_fn_t fn = find_close_rdma_session();
+    if (!fn) {
+        // Symbol not exported — session cleanup handled by destructor.
+        return;
+    }
+    fn(reinterpret_cast<void *>(srv));
+}
+
+int cuobj_server_is_connected(cuobj_server_t *srv) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    return s->isConnected() ? 1 : 0;
+}
+
+void* cuobj_server_alloc_host_buffer(cuobj_server_t *srv, size_t size) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    return s->allocHostBuffer(size);
+}
+
+void cuobj_server_free_host_buffer(void *ptr) {
+    free(ptr);
+}
+
+cuobj_rdma_buffer_t* cuobj_server_register_buffer(cuobj_server_t *srv, void *ptr, size_t size) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    return s->registerBuffer(ptr, size);
+}
+
+void cuobj_server_deregister_buffer(cuobj_server_t *srv, cuobj_rdma_buffer_t *buf) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    s->deRegisterBuffer(buf);
+}
+
+uint16_t cuobj_server_allocate_channel(cuobj_server_t *srv) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    return s->allocateChannelId();
+}
+
+void cuobj_server_free_channel(cuobj_server_t *srv, uint16_t channel_id) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    s->freeChannelId(channel_id);
+}
+
+ssize_t cuobj_server_handle_get(cuobj_server_t *srv,
+                                const char *key,
+                                cuobj_rdma_buffer_t *local_buf,
+                                uint64_t remote_buf_start,
+                                size_t size,
+                                const char *rdma_descr,
+                                uint16_t channel) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    std::string k(key);
+    std::string descr(rdma_descr);
+    ibv_wc_status wc_status = IBV_WC_SUCCESS;
+    ssize_t rc = s->handleGetObject(k, local_buf, remote_buf_start, size, descr,
+                                    channel, 0, &wc_status);
+    if (rc < 0 && g_verbose_logs.load()) {
+        fprintf(stderr,
+                "cuobjwrapper: handleGetObject rc=%zd channel=%u wc_status=%d\n",
+                rc, static_cast<unsigned>(channel), static_cast<int>(wc_status));
+    }
+    return rc;
+}
+
+ssize_t cuobj_server_handle_put(cuobj_server_t *srv,
+                                const char *key,
+                                cuobj_rdma_buffer_t *local_buf,
+                                uint64_t remote_buf_start,
+                                size_t size,
+                                const char *rdma_descr,
+                                uint16_t channel) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    std::string k(key);
+    std::string descr(rdma_descr);
+    ibv_wc_status wc_status = IBV_WC_SUCCESS;
+    ssize_t rc = s->handlePutObject(k, local_buf, remote_buf_start, size, descr,
+                                    channel, 0, &wc_status);
+    if (rc < 0 && g_verbose_logs.load()) {
+        fprintf(stderr,
+                "cuobjwrapper: handlePutObject rc=%zd channel=%u wc_status=%d\n",
+                rc, static_cast<unsigned>(channel), static_cast<int>(wc_status));
+    }
+    return rc;
+}
+
+void cuobj_server_setup_telemetry(int use_otel) {
+    cuObjServer::setupTelemetry(use_otel != 0, &std::cout);
+}
+
+void cuobj_server_shutdown_telemetry(void) {
+    cuObjServer::shutdownTelemetry();
+}
+
+void cuobj_server_set_telem_flags(unsigned flags) {
+    g_verbose_logs.store((flags & 0x0003u) != 0u);
+    cuObjServer::setTelemFlags(flags);
+}
+
+int cuobj_server_init_rdma_config(cuobj_server_t *srv, const cuobj_rdma_tunables_t *t) {
+    auto *s = reinterpret_cast<cuObjServer*>(srv);
+    try {
+        s->initRDMAConfigParams(tunables_from_c(t));
+        return 0;
+    } catch (...) {
+        return -1;
+    }
+}
+
+} // extern "C"
@@ -0,0 +1,117 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// C wrapper for the cuObjServer C++ class.
+// Provides a C-linkage interface suitable for CGO consumption.
+
+#ifndef CUOBJSERVER_WRAPPER_H
+#define CUOBJSERVER_WRAPPER_H
+
+#include <stddef.h>
+#include <stdint.h>
+#include <sys/types.h>
+
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+// Opaque handles
+typedef struct cuobj_server cuobj_server_t;
+typedef struct rdma_buffer cuobj_rdma_buffer_t;
+
+// Error codes matching cuObjErr_t
+#define CUOBJ_OK   0
+#define CUOBJ_FAIL 1
+
+// cuobj_server lifecycle
+cuobj_server_t* cuobj_server_create(const char *ip, unsigned short port, unsigned proto);
+void            cuobj_server_destroy(cuobj_server_t *srv);
+
+// RDMA session
+int  cuobj_server_start_session(cuobj_server_t *srv);
+void cuobj_server_close_session(cuobj_server_t *srv);
+int  cuobj_server_is_connected(cuobj_server_t *srv);
+
+// Host buffer allocation
+void* cuobj_server_alloc_host_buffer(cuobj_server_t *srv, size_t size);
+void  cuobj_server_free_host_buffer(void *ptr);
+
+// Buffer registration
+cuobj_rdma_buffer_t* cuobj_server_register_buffer(cuobj_server_t *srv, void *ptr, size_t size);
+void                 cuobj_server_deregister_buffer(cuobj_server_t *srv, cuobj_rdma_buffer_t *buf);
+
+// Channel management
+uint16_t cuobj_server_allocate_channel(cuobj_server_t *srv);
+void     cuobj_server_free_channel(cuobj_server_t *srv, uint16_t channel_id);
+
+// Data transfer (synchronous, no poll_delay override)
+//
+// handleGetObject: RDMA WRITE server→client (serves a GET request)
+//   Returns bytes transferred or -1 on error.
+ssize_t cuobj_server_handle_get(cuobj_server_t *srv,
+                                const char *key,
+                                cuobj_rdma_buffer_t *local_buf,
+                                uint64_t remote_buf_start,
+                                size_t size,
+                                const char *rdma_descr,
+                                uint16_t channel);
+
+// handlePutObject: RDMA READ client→server (serves a PUT request)
+//   Returns bytes transferred or -1 on error.
+ssize_t cuobj_server_handle_put(cuobj_server_t *srv,
+                                const char *key,
+                                cuobj_rdma_buffer_t *local_buf,
+                                uint64_t remote_buf_start,
+                                size_t size,
+                                const char *rdma_descr,
+                                uint16_t channel);
+
+// Telemetry (optional)
+void cuobj_server_setup_telemetry(int use_otel);
+void cuobj_server_shutdown_telemetry(void);
+void cuobj_server_set_telem_flags(unsigned flags);
+
+// RDMA tunable parameters — flat C struct for CGO compatibility.
+// Field names and defaults match cuObjRDMATunableParam in cuobjrdma.h.
+typedef struct {
+    int           num_dcis;             // default 128
+    unsigned      cq_depth;             // default 640
+    unsigned long dc_key;               // default 0xffeeddcc
+    int           service_level;        // default 0
+    uint8_t       timeout;              // default 16
+    unsigned      hop_limit;            // default 4
+    int           pkey_index;           // default 0
+    uint32_t      delay_interval;       // default 5000 ns
+    int           delay_mode;           // 0=none 1=batch 2=entry 3=adaptive; default 1
+    uint8_t       retry_cnt;            // default 7
+    int           qp_reset_on_failure;  // bool as int; default 1 (true)
+    unsigned      traffic_class;        // default 96
+    int           max_rd_atomic;        // default 0 (auto)
+} cuobj_rdma_tunables_t;
+
+// Apply RDMA tuning parameters to an existing connection object.
+// Takes effect on the next reconnection if called after session start.
+// Returns 0 on success, -1 on error.
+int cuobj_server_init_rdma_config(cuobj_server_t *srv, const cuobj_rdma_tunables_t *t);
+
+// Create a cuObjServer with tunable parameters applied before the session
+// starts. This is the preferred constructor when non-default tunables are
+// needed, since the library starts the RDMA session inside the constructor.
+cuobj_server_t* cuobj_server_create_with_config(const char *ip, unsigned short port, unsigned proto, const cuobj_rdma_tunables_t *t);
+
+#ifdef __cplusplus
+}
+#endif
+
+#endif // CUOBJSERVER_WRAPPER_H
@@ -0,0 +1,460 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// Host-memory RDMA client wrapper implementation.
+//
+// Builds a passive DC Target (DCT) endpoint with libibverbs + mlx5 direct
+// verbs so a cuObjServer can RDMA READ/WRITE the client's registered host
+// memory. No CUDA/GPU dependency.
+
+#include "rdma_host_client_wrapper.h"
+
+#include <cstdio>
+#include <cstdlib>
+#include <cstring>
+#include <limits>
+#include <new>
+#include <string>
+
+#include <infiniband/verbs.h>
+#include <infiniband/mlx5dv.h>
+
+struct rdma_host_client {
+    struct ibv_context *ctx = nullptr;
+    struct ibv_pd      *pd  = nullptr;
+    struct ibv_cq      *cq  = nullptr;
+    struct ibv_srq     *srq = nullptr;
+    struct ibv_qp      *dct = nullptr; // DC target QP
+
+    uint8_t  port_num  = 1;
+    int      gid_index = 0;
+    uint64_t dc_key    = 0;
+
+    uint16_t     lid   = 0;
+    union ibv_gid gid  = {};
+    uint32_t     dctn  = 0;
+
+    void       *buf     = nullptr;
+    size_t      buf_len = 0;
+    struct ibv_mr *mr   = nullptr;
+
+    std::string token;
+    std::string last_error;
+};
+
+static void set_err(rdma_host_client_t *c, const std::string &msg) {
+    if (c) {
+        c->last_error = msg;
+    }
+}
+
+static bool gid_is_zero(const union ibv_gid &gid) {
+    for (int i = 0; i < 16; i++) {
+        if (gid.raw[i] != 0) {
+            return false;
+        }
+    }
+    return true;
+}
+
+static bool gid_is_link_local(const union ibv_gid &gid) {
+    return gid.raw[0] == 0xfe && (gid.raw[1] & 0xc0) == 0x80;
+}
+
+static bool gid_is_ipv4_mapped(const union ibv_gid &gid) {
+    for (int i = 0; i < 10; i++) {
+        if (gid.raw[i] != 0) {
+            return false;
+        }
+    }
+    return gid.raw[10] == 0xff && gid.raw[11] == 0xff;
+}
+
+// gid_is_roce_v2 reports whether GID index i is a RoCEv2 entry. mlx5
+// commonly exposes a RoCEv1 and RoCEv2 GID pair with identical raw bytes at
+// adjacent indices (v1 first), so the raw GID content alone cannot
+// distinguish them; only ibv_query_gid_ex()'s reported gid_type can. If the
+// query fails (e.g. native InfiniBand link layer, where the v1/v2
+// distinction doesn't apply), the GID is not excluded on this basis.
+static bool gid_is_roce_v2(struct ibv_context *ctx, uint8_t port_num, int index) {
+    struct ibv_gid_entry entry;
+    memset(&entry, 0, sizeof(entry));
+    if (ibv_query_gid_ex(ctx, port_num, static_cast<uint32_t>(index), &entry, 0)) {
+        return true;
+    }
+    return entry.gid_type == IBV_GID_TYPE_ROCE_V2;
+}
+
+static bool select_gid(rdma_host_client_t *c) {
+    if (c->gid_index >= 0) {
+        if (ibv_query_gid(c->ctx, c->port_num, c->gid_index, &c->gid)) {
+            set_err(c, "ibv_query_gid failed");
+            return false;
+        }
+        return true;
+    }
+
+    struct ibv_port_attr port_attr;
+    memset(&port_attr, 0, sizeof(port_attr));
+    if (ibv_query_port(c->ctx, c->port_num, &port_attr)) {
+        set_err(c, "ibv_query_port failed while selecting GID");
+        return false;
+    }
+
+    int fallback_gid_index = -1;
+    union ibv_gid fallback_gid = {};
+    int v1_fallback_gid_index = -1;
+    union ibv_gid v1_fallback_gid = {};
+    int v1_non_ipv4_gid_index = -1;
+    union ibv_gid v1_non_ipv4_gid = {};
+
+    // Prefer non-zero, non-link-local, non-IPv4-mapped, RoCEv2 GIDs.
+    // Use ibv_query_gid() only for broad provider compatibility.
+    for (int i = 0; i < port_attr.gid_tbl_len; i++) {
+        union ibv_gid gid = {};
+        if (ibv_query_gid(c->ctx, c->port_num, i, &gid)) {
+            continue;
+        }
+        if (gid_is_zero(gid)) {
+            continue;
+        }
+        if (gid_is_link_local(gid)) {
+            continue;
+        }
+        bool is_v2 = gid_is_roce_v2(c->ctx, c->port_num, i);
+        if (gid_is_ipv4_mapped(gid)) {
+            if (is_v2) {
+                if (fallback_gid_index < 0) {
+                    fallback_gid_index = i;
+                    fallback_gid = gid;
+                }
+            } else if (v1_fallback_gid_index < 0) {
+                v1_fallback_gid_index = i;
+                v1_fallback_gid = gid;
+            }
+            continue;
+        }
+        if (is_v2) {
+            c->gid_index = i;
+            c->gid = gid;
+            return true;
+        }
+        if (v1_non_ipv4_gid_index < 0) {
+            v1_non_ipv4_gid_index = i;
+            v1_non_ipv4_gid = gid;
+        }
+    }
+
+    // Fallbacks, in order of preference: RoCEv2 IPv4-mapped, then any
+    // RoCEv1 candidate. RoCEv1 is only used if no RoCEv2 GID is present at
+    // all (e.g. non-Mellanox providers where ibv_query_gid_ex reports
+    // unknown type for every entry, in which case gid_is_roce_v2 treats all
+    // entries as usable and the earlier non-IPv4-mapped branch above already
+    // returns on the first candidate).
+    if (fallback_gid_index >= 0) {
+        c->gid_index = fallback_gid_index;
+        c->gid = fallback_gid;
+        return true;
+    }
+    if (v1_non_ipv4_gid_index >= 0) {
+        c->gid_index = v1_non_ipv4_gid_index;
+        c->gid = v1_non_ipv4_gid;
+        return true;
+    }
+    if (v1_fallback_gid_index >= 0) {
+        c->gid_index = v1_fallback_gid_index;
+        c->gid = v1_fallback_gid;
+        return true;
+    }
+
+    set_err(c,
+            "no usable GID found on selected RDMA port; "
+            "set VGWRDMA_GID_INDEX to a valid RoCE GID index from `ibv_devinfo -v`");
+    return false;
+}
+
+// build_token encodes the RDMA descriptor for the currently registered region.
+//
+// IMPORTANT: the exact on-wire token format is defined by the NVIDIA cuObject
+// RDMA DC protocol as parsed by libcuobjserver. It is a colon-delimited,
+// lowercase-hex string whose 7 fields are, in order:
+//
+//   # | Field                                     | Type     | Width
+//   --|--------------------------------------------|----------|-----------
+//   1 | Remote base address (GPUMEM/SYSMEM)         | uint64   | 16 chars
+//   2 | Max size of buffer region from base addr    | uint32   | 8 chars
+//   3 | Remote key (rkey)                           | uint32   | 8 chars
+//   4 | LID of the client NIC                       | uint16   | 4 chars
+//   5 | DCTN                                        | uint32   | 6 chars
+//   6 | GID present (1|0)                           | bool     | 1 char
+//   7 | GID of client NIC                           | 16 bytes | 32 chars
+//
+// Example: "0102030405060708:01020304:01020304:0102:010203:1:0102030405060708090a0b0c0d0e0f10"
+//
+// The canonical definition (consumed by the Go gateway) lives in
+// cumiddleware/cuobj.go next to HeaderRDMAToken; keep the two in sync.
+// Assumptions kept explicit here:
+//   1) Field widths and lowercase hex formatting are strict parser contracts.
+//   2) GID is serialized as the raw 16 bytes in order (no hextet splitting).
+//   3) Descriptor size field is uint32; oversized buffers are rejected.
+static void build_token(rdma_host_client_t *c) {
+    char gid_hex[33];
+    for (int i = 0; i < 16; i++) {
+        snprintf(&gid_hex[i * 2], 3, "%02x", c->gid.raw[i]);
+    }
+    gid_hex[32] = '\0';
+
+    char buf[320];
+    snprintf(buf, sizeof(buf),
+             "%016llx:%08x:%08x:%04x:%06x:%1x:%s",
+             (unsigned long long)(uintptr_t)c->buf,
+             static_cast<unsigned>(c->buf_len),
+             c->mr ? c->mr->rkey : 0u,
+             static_cast<unsigned>(c->lid),
+             static_cast<unsigned>(c->dctn),
+             1, // GID present (RoCE)
+             gid_hex);
+    c->token = buf;
+}
+
+static bool build_dct(rdma_host_client_t *c) {
+    c->cq = ibv_create_cq(c->ctx, 1, nullptr, nullptr, 0);
+    if (!c->cq) {
+        set_err(c, "ibv_create_cq failed");
+        return false;
+    }
+
+    struct ibv_srq_init_attr srq_attr;
+    memset(&srq_attr, 0, sizeof(srq_attr));
+    srq_attr.attr.max_wr  = 1;
+    srq_attr.attr.max_sge = 1;
+    c->srq = ibv_create_srq(c->pd, &srq_attr);
+    if (!c->srq) {
+        set_err(c, "ibv_create_srq failed");
+        return false;
+    }
+
+    struct ibv_qp_init_attr_ex attr_ex;
+    memset(&attr_ex, 0, sizeof(attr_ex));
+    attr_ex.pd        = c->pd;
+    attr_ex.send_cq   = c->cq;
+    attr_ex.recv_cq   = c->cq;
+    attr_ex.srq       = c->srq;
+    attr_ex.qp_type   = IBV_QPT_DRIVER;
+    attr_ex.comp_mask = IBV_QP_INIT_ATTR_PD;
+
+    struct mlx5dv_qp_init_attr dv_attr;
+    memset(&dv_attr, 0, sizeof(dv_attr));
+    dv_attr.comp_mask = MLX5DV_QP_INIT_ATTR_MASK_DC;
+    dv_attr.dc_init_attr.dc_type         = MLX5DV_DCTYPE_DCT;
+    dv_attr.dc_init_attr.dct_access_key  = c->dc_key;
+
+    c->dct = mlx5dv_create_qp(c->ctx, &attr_ex, &dv_attr);
+    if (!c->dct) {
+        set_err(c, "mlx5dv_create_qp (DCT) failed");
+        return false;
+    }
+
+    // INIT
+    struct ibv_qp_attr qpa;
+    memset(&qpa, 0, sizeof(qpa));
+    qpa.qp_state        = IBV_QPS_INIT;
+    qpa.pkey_index      = 0;
+    qpa.port_num        = c->port_num;
+    qpa.qp_access_flags = IBV_ACCESS_REMOTE_WRITE | IBV_ACCESS_REMOTE_READ;
+    if (ibv_modify_qp(c->dct, &qpa,
+                      IBV_QP_STATE | IBV_QP_PKEY_INDEX | IBV_QP_PORT |
+                      IBV_QP_ACCESS_FLAGS)) {
+        set_err(c, "ibv_modify_qp DCT->INIT failed");
+        return false;
+    }
+
+    struct ibv_port_attr port_attr;
+    memset(&port_attr, 0, sizeof(port_attr));
+    if (ibv_query_port(c->ctx, c->port_num, &port_attr)) {
+        set_err(c, "ibv_query_port failed");
+        return false;
+    }
+    c->lid = port_attr.lid;
+
+    // RTR — DCT only needs INIT->RTR (no RTS for targets).
+    memset(&qpa, 0, sizeof(qpa));
+    qpa.qp_state           = IBV_QPS_RTR;
+    qpa.path_mtu           = port_attr.active_mtu;
+    qpa.min_rnr_timer      = 12;
+    qpa.ah_attr.is_global  = 1;
+    qpa.ah_attr.port_num   = c->port_num;
+    qpa.ah_attr.grh.hop_limit     = 4;
+    qpa.ah_attr.grh.sgid_index    = (uint8_t)c->gid_index;
+    qpa.ah_attr.grh.traffic_class = 0;
+    if (ibv_modify_qp(c->dct, &qpa,
+                      IBV_QP_STATE | IBV_QP_MIN_RNR_TIMER | IBV_QP_AV |
+                      IBV_QP_PATH_MTU)) {
+        set_err(c, "ibv_modify_qp DCT->RTR failed");
+        return false;
+    }
+
+    // Read DCT number after INIT->RTR is complete. Some provider paths may
+    // expose qp_num late; encoding 0 here breaks remote addressing.
+    c->dctn = c->dct->qp_num;
+    if (c->dctn == 0) {
+        set_err(c, "DCT number is zero after RTR transition");
+        return false;
+    }
+
+    return true;
+}
+
+extern "C" {
+
+rdma_host_client_t* rdma_host_client_create(const char *dev_name,
+                                            uint8_t port_num,
+                                            int gid_index,
+                                            uint64_t dc_key) {
+    auto *c = new (std::nothrow) rdma_host_client();
+    if (!c) {
+        return nullptr;
+    }
+    c->port_num  = port_num ? port_num : 1;
+    c->gid_index = gid_index;
+    c->dc_key    = dc_key;
+
+    int num = 0;
+    struct ibv_device **list = ibv_get_device_list(&num);
+    if (!list || num == 0) {
+        set_err(c, "ibv_get_device_list found no devices");
+        if (list) ibv_free_device_list(list);
+        return c; // return handle so caller can read last_error
+    }
+
+    struct ibv_device *dev = nullptr;
+    if (dev_name && dev_name[0]) {
+        for (int i = 0; i < num; i++) {
+            if (strcmp(ibv_get_device_name(list[i]), dev_name) == 0) {
+                dev = list[i];
+                break;
+            }
+        }
+        if (!dev) {
+            set_err(c, std::string("RDMA device not found: ") + dev_name);
+            ibv_free_device_list(list);
+            return c;
+        }
+    } else {
+        dev = list[0];
+    }
+
+    c->ctx = ibv_open_device(dev);
+    ibv_free_device_list(list);
+    if (!c->ctx) {
+        set_err(c, "ibv_open_device failed");
+        return c;
+    }
+
+    c->pd = ibv_alloc_pd(c->ctx);
+    if (!c->pd) {
+        set_err(c, "ibv_alloc_pd failed");
+        return c;
+    }
+
+    if (!select_gid(c)) {
+        return c;
+    }
+
+    if (!build_dct(c)) {
+        return c;
+    }
+
+    c->last_error.clear();
+    return c;
+}
+
+void rdma_host_client_free(rdma_host_client_t *c) {
+    if (!c) {
+        return;
+    }
+    if (c->mr) {
+        ibv_dereg_mr(c->mr);
+        c->mr = nullptr;
+    }
+    if (c->buf) {
+        free(c->buf);
+        c->buf = nullptr;
+    }
+    c->buf_len = 0;
+    c->token.clear();
+}
+
+void* rdma_host_client_alloc(rdma_host_client_t *c, size_t size) {
+    if (!c || size == 0) {
+        return nullptr;
+    }
+    if (size > static_cast<size_t>(std::numeric_limits<uint32_t>::max())) {
+        set_err(c, "requested buffer size exceeds descriptor uint32 max");
+        return nullptr;
+    }
+    rdma_host_client_free(c);
+
+    void *ptr = nullptr;
+    if (posix_memalign(&ptr, 4096, size) != 0 || !ptr) {
+        set_err(c, "posix_memalign failed");
+        return nullptr;
+    }
+    memset(ptr, 0, size);
+
+    struct ibv_mr *mr = ibv_reg_mr(c->pd, ptr, size,
+                                   IBV_ACCESS_LOCAL_WRITE |
+                                   IBV_ACCESS_REMOTE_WRITE |
+                                   IBV_ACCESS_REMOTE_READ);
+    if (!mr) {
+        set_err(c, "ibv_reg_mr failed");
+        free(ptr);
+        return nullptr;
+    }
+
+    c->buf     = ptr;
+    c->buf_len = size;
+    c->mr      = mr;
+    build_token(c);
+    return ptr;
+}
+
+const char* rdma_host_client_token(rdma_host_client_t *c) {
+    if (!c || c->token.empty()) {
+        return nullptr;
+    }
+    return c->token.c_str();
+}
+
+const char* rdma_host_client_last_error(rdma_host_client_t *c) {
+    if (!c || c->last_error.empty()) {
+        return nullptr;
+    }
+    return c->last_error.c_str();
+}
+
+void rdma_host_client_destroy(rdma_host_client_t *c) {
+    if (!c) {
+        return;
+    }
+    rdma_host_client_free(c);
+    if (c->dct) ibv_destroy_qp(c->dct);
+    if (c->srq) ibv_destroy_srq(c->srq);
+    if (c->cq)  ibv_destroy_cq(c->cq);
+    if (c->pd)  ibv_dealloc_pd(c->pd);
+    if (c->ctx) ibv_close_device(c->ctx);
+    delete c;
+}
+
+} // extern "C"
@@ -0,0 +1,80 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// Host-memory RDMA client wrapper.
+//
+// This wrapper implements the *client* side of the cuObject RDMA DC (Dynamically
+// Connected) protocol using libibverbs + mlx5 direct-verbs, WITHOUT any CUDA/GPU
+// dependency. It is intended for RDMA-capable hosts that have no GPU and cannot
+// use the official NVIDIA libcuobjclient library (which requires CUDA).
+//
+// Role in the protocol: the cuObjServer (gateway) is the RDMA *initiator*
+// (it owns the DCI QPs and issues RDMA READ for PUT / RDMA WRITE for GET).
+// The client is the passive *target*: it exposes a DC Target (DCT) QP plus a
+// registered memory region, and encodes their coordinates into an RDMA
+// descriptor token that travels to the server in an S3 request header. During
+// the actual data transfer the client CPU is not involved — the NIC services
+// the server's reads/writes against the registered region via the DCT.
+//
+// The C ABI below is CGO-friendly (opaque handle, C linkage, no C++ types).
+
+#ifndef RDMA_HOST_CLIENT_WRAPPER_H
+#define RDMA_HOST_CLIENT_WRAPPER_H
+
+#include <stddef.h>
+#include <stdint.h>
+
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+typedef struct rdma_host_client rdma_host_client_t;
+
+// rdma_host_client_create opens the given RDMA device and builds the passive
+// DC target endpoint (PD, CQ, SRQ, DCT QP -> RTR) plus caches the port LID and
+// the GID at gid_index. dc_key must match the server's DCKey tunable
+// (default 0xffeeddcc). Pass dev_name = NULL to select the first device.
+// Returns NULL on failure.
+rdma_host_client_t* rdma_host_client_create(const char *dev_name,
+                                            uint8_t port_num,
+                                            int gid_index,
+                                            uint64_t dc_key);
+
+// rdma_host_client_destroy tears down the endpoint and frees any registered
+// buffer. Safe to call with NULL.
+void rdma_host_client_destroy(rdma_host_client_t *c);
+
+// rdma_host_client_alloc allocates a page-aligned host buffer of size bytes and
+// registers it for remote read+write. Any previously allocated buffer is freed
+// first. Returns the buffer pointer, or NULL on failure.
+void* rdma_host_client_alloc(rdma_host_client_t *c, size_t size);
+
+// rdma_host_client_free deregisters and frees the current buffer, if any.
+void rdma_host_client_free(rdma_host_client_t *c);
+
+// rdma_host_client_token returns the RDMA descriptor token for the currently
+// registered buffer, or NULL if no buffer is registered. The returned string is
+// owned by the client and remains valid until the next alloc/free/destroy.
+// See build_token() in rdma_host_client_wrapper.cpp for the wire-format field table.
+const char* rdma_host_client_token(rdma_host_client_t *c);
+
+// rdma_host_client_last_error returns a human-readable description of the most
+// recent failure, or NULL if none.
+const char* rdma_host_client_last_error(rdma_host_client_t *c);
+
+#ifdef __cplusplus
+}
+#endif
+
+#endif // RDMA_HOST_CLIENT_WRAPPER_H
@@ -0,0 +1,98 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+// Package bufferpool provides a pool of pre-registered RDMA buffers
+// to avoid per-request allocation and registration overhead.
+package bufferpool
+
+import (
+	"context"
+	"errors"
+	"fmt"
+
+	"github.com/versity/versitygw/rdma"
+)
+
+// Pool manages a fixed set of RDMA-registered host memory buffers.
+type Pool struct {
+	server  *rdma.Server
+	bufSize int
+	ch      chan *rdma.Buffer
+	all     []*rdma.Buffer
+}
+
+// NewPool pre-allocates and registers count buffers of bufSize bytes each.
+func NewPool(server *rdma.Server, bufSize, count int) (*Pool, error) {
+	if count <= 0 {
+		return nil, errors.New("bufferpool: count must be > 0")
+	}
+	if bufSize <= 0 {
+		return nil, errors.New("bufferpool: bufSize must be > 0")
+	}
+
+	p := &Pool{
+		server:  server,
+		bufSize: bufSize,
+		ch:      make(chan *rdma.Buffer, count),
+		all:     make([]*rdma.Buffer, 0, count),
+	}
+
+	for i := range count {
+		ptr, err := server.AllocHostBuffer(bufSize)
+		if err != nil {
+			p.Close()
+			return nil, fmt.Errorf("bufferpool: alloc buffer %d: %w", i, err)
+		}
+
+		buf, err := server.RegisterBuffer(ptr, bufSize)
+		if err != nil {
+			server.FreeHostBuffer(ptr)
+			p.Close()
+			return nil, fmt.Errorf("bufferpool: register buffer %d: %w", i, err)
+		}
+
+		p.all = append(p.all, buf)
+		p.ch <- buf
+	}
+
+	return p, nil
+}
+
+// Acquire blocks until a buffer is available or ctx is cancelled.
+// Returns the buffer and its backing byte slice.
+func (p *Pool) Acquire(ctx context.Context) (*rdma.Buffer, []byte, error) {
+	select {
+	case buf := <-p.ch:
+		return buf, buf.Slice(), nil
+	case <-ctx.Done():
+		return nil, nil, ctx.Err()
+	}
+}
+
+// Release returns a buffer to the pool.
+func (p *Pool) Release(buf *rdma.Buffer) {
+	p.ch <- buf
+}
+
+// Close deregisters and frees all buffers. The pool must not be used afterward.
+func (p *Pool) Close() {
+	for _, buf := range p.all {
+		// Capture the host pointer before DeregisterBuffer clears it, so the
+		// underlying allocation from AllocHostBuffer can still be freed.
+		ptr := buf.HostPtr()
+		p.server.DeregisterBuffer(buf)
+		p.server.FreeHostBuffer(ptr)
+	}
+	p.all = nil
+}
@@ -0,0 +1,143 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build linux && amd64 && cgo
+
+// Package hostclient implements the client side of the cuObject RDMA DC
+// protocol using libibverbs + mlx5 direct verbs, with no CUDA/GPU dependency.
+//
+// It is the host-memory counterpart to the GPU-based NVIDIA libcuobjclient:
+// it registers host memory, exposes a passive DC target the gateway can RDMA
+// READ/WRITE, and produces the RDMA descriptor token the gateway expects in
+// the S3 request headers.
+package hostclient
+
+/*
+#cgo CFLAGS: -I${SRCDIR}/../../cuwrapper
+#cgo LDFLAGS: -L${SRCDIR}/.. -l:libhostclientwrapper.a -libverbs -lmlx5 -lstdc++
+#include <stdlib.h>
+#include "rdma_host_client_wrapper.h"
+*/
+import "C"
+
+import (
+	"errors"
+	"fmt"
+	"unsafe"
+)
+
+// DefaultDCKey is the cuObjServer default Dynamic Connection key. The client
+// DC target must use a key matching the server's DCKey tunable.
+const DefaultDCKey uint64 = 0xffeeddcc
+
+// Client owns a passive DC target endpoint and a single registered host buffer.
+// A Client is not safe for concurrent use.
+type Client struct {
+	c    *C.rdma_host_client_t
+	buf  unsafe.Pointer
+	size int
+}
+
+// NewClient opens the RDMA device and builds the DC target endpoint.
+// dev selects the device by name (e.g. "mlx5_0"); empty picks the first.
+// port is the HCA port (usually 1). gidIndex selects the RoCE GID (see
+// `ibv_devinfo -v`). dcKey must match the server's DCKey (DefaultDCKey by
+// default).
+func NewClient(dev string, port uint8, gidIndex int, dcKey uint64) (*Client, error) {
+	var cdev *C.char
+	if dev != "" {
+		cdev = C.CString(dev)
+		defer C.free(unsafe.Pointer(cdev))
+	}
+	h := C.rdma_host_client_create(cdev, C.uint8_t(port), C.int(gidIndex), C.uint64_t(dcKey))
+	if h == nil {
+		return nil, errors.New("hostclient: allocation failed")
+	}
+	c := &Client{c: h}
+	if msg := C.rdma_host_client_last_error(h); msg != nil {
+		err := fmt.Errorf("hostclient: %s", C.GoString(msg))
+		c.Close()
+		return nil, err
+	}
+	return c, nil
+}
+
+// Register allocates and registers a host buffer of the given size and returns
+// a Go slice backed by it. The slice is valid until the next Register call or
+// Close. Copy PUT data into the slice before the transfer; read GET data out
+// of it afterward.
+func (c *Client) Register(size int) ([]byte, error) {
+	if c.c == nil {
+		return nil, errors.New("hostclient: closed")
+	}
+	if size <= 0 {
+		return nil, fmt.Errorf("hostclient: invalid size %d", size)
+	}
+	ptr := C.rdma_host_client_alloc(c.c, C.size_t(size))
+	if ptr == nil {
+		return nil, fmt.Errorf("hostclient: register failed: %s", c.lastError())
+	}
+	c.buf = unsafe.Pointer(ptr)
+	c.size = size
+	return unsafe.Slice((*byte)(ptr), size), nil
+}
+
+// Token returns the RDMA descriptor token for the currently registered buffer.
+// See cumiddleware.HeaderRDMAToken for the canonical wire-format field table.
+func (c *Client) Token() (string, error) {
+	if c.c == nil {
+		return "", errors.New("hostclient: closed")
+	}
+	t := C.rdma_host_client_token(c.c)
+	if t == nil {
+		return "", errors.New("hostclient: no registered buffer")
+	}
+	return C.GoString(t), nil
+}
+
+// Buffer returns the current registered buffer as a Go slice, or nil.
+func (c *Client) Buffer() []byte {
+	if c.c == nil || c.buf == nil || c.size == 0 {
+		return nil
+	}
+	return unsafe.Slice((*byte)(c.buf), c.size)
+}
+
+// BufferAddr returns the virtual address of the registered buffer. This is the
+// remote base address the gateway uses as the RDMA start offset; it matches the
+// address encoded in Token.
+func (c *Client) BufferAddr() uint64 {
+	if c.c == nil || c.buf == nil {
+		return 0
+	}
+	return uint64(uintptr(c.buf))
+}
+
+// Close releases the buffer and tears down the endpoint.
+func (c *Client) Close() {
+	if c.c == nil {
+		return
+	}
+	C.rdma_host_client_destroy(c.c)
+	c.c = nil
+	c.buf = nil
+	c.size = 0
+}
+
+func (c *Client) lastError() string {
+	if msg := C.rdma_host_client_last_error(c.c); msg != nil {
+		return C.GoString(msg)
+	}
+	return "unknown error"
+}
@@ -0,0 +1,48 @@
+// Copyright 2026 Versity Software
+// This file is licensed under the Apache License, Version 2.0
+// (the "License"); you may not use this file except in compliance
+// with the License.  You may obtain a copy of the License at
+//
+//   http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing,
+// software distributed under the License is distributed on an
+// "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+// KIND, either express or implied.  See the License for the
+// specific language governing permissions and limitations
+// under the License.
+
+//go:build !(linux && amd64 && cgo)
+
+// Package hostclient is stubbed on platforms without RDMA verbs support.
+package hostclient
+
+import "errors"
+
+// DefaultDCKey is the cuObjServer default Dynamic Connection key.
+const DefaultDCKey uint64 = 0xffeeddcc
+
+var errNotSupported = errors.New("hostclient: RDMA host client is only supported on linux/amd64 with cgo")
+
+// Client is a non-functional stub on unsupported platforms.
+type Client struct{}
+
+// NewClient always returns an unsupported-platform error on this build.
+func NewClient(dev string, port uint8, gidIndex int, dcKey uint64) (*Client, error) {
+	return nil, errNotSupported
+}
+
+// Register always returns an unsupported-platform error on this build.
+func (c *Client) Register(size int) ([]byte, error) { return nil, errNotSupported }
+
+// Token always returns an unsupported-platform error on this build.
+func (c *Client) Token() (string, error) { return "", errNotSupported }
+
+// Buffer returns nil on this build.
+func (c *Client) Buffer() []byte { return nil }
+
+// BufferAddr returns 0 on this build.
+func (c *Client) BufferAddr() uint64 { return 0 }
+
+// Close is a no-op on this build.
+func (c *Client) Close() {}
@@ -0,0 +1,298 @@
+//go:build linux && amd64 && cgo
+
+// Package rdma provides Go bindings to libcuobjserver via CGO.
+package rdma
+
+/*
+#cgo CFLAGS: -I${SRCDIR}/../include -I${SRCDIR}/../cuwrapper
+#cgo LDFLAGS: -L${SRCDIR} -l:libcuobjwrapper.a -L/usr/lib64 -lcuobjserver -lstdc++ -ldl
+#include "cuobjserver_wrapper.h"
+#include <stdlib.h>
+*/
+import "C"
+
+import (
+	"errors"
+	"fmt"
+	"sync"
+	"sync/atomic"
+	"unsafe"
+)
+
+const (
+	cuobjLogPathInfo  = 0x0001
+	cuobjLogPathDebug = 0x0002
+	cuobjLogPathError = 0x0004
+
+	// cuobjProtoRDMADCV1 is the CUOBJ_PROTO_RDMA_DC_V1 protocol identifier.
+	cuobjProtoRDMADCV1 = 1001
+)
+
+var debugTelemetryEnabled atomic.Bool
+
+// ConfigureTelemetry controls cuObjServer telemetry logging.
+// When debug is true, enable info+debug+error logs to stderr/stdout path.
+func ConfigureTelemetry(debug bool) {
+	debugTelemetryEnabled.Store(debug)
+	C.cuobj_server_setup_telemetry(0)
+	if debug {
+		C.cuobj_server_set_telem_flags(C.uint(cuobjLogPathInfo | cuobjLogPathDebug | cuobjLogPathError))
+		return
+	}
+	C.cuobj_server_set_telem_flags(C.uint(cuobjLogPathError))
+}
+
+// DebugTelemetryEnabled reports whether verbose RDMA diagnostics are enabled.
+func DebugTelemetryEnabled() bool {
+	return debugTelemetryEnabled.Load()
+}
+
+// Buffer wraps an RDMA-registered memory region.
+type Buffer struct {
+	cbuf    *C.cuobj_rdma_buffer_t
+	hostPtr unsafe.Pointer
+	size    int
+}
+
+// HostPtr returns the underlying host memory pointer.
+func (b *Buffer) HostPtr() unsafe.Pointer { return b.hostPtr }
+
+// Size returns the buffer size in bytes.
+func (b *Buffer) Size() int { return b.size }
+
+// Slice returns the buffer contents as a Go byte slice backed by C-allocated
+// memory. Returns nil after DeregisterBuffer.
+func (b *Buffer) Slice() []byte {
+	return unsafe.Slice((*byte)(b.hostPtr), b.size)
+}
+
+// Server wraps a cuObjServer instance.
+type Server struct {
+	csrv        *C.cuobj_server_t
+	sessionOpen bool
+	mu          sync.Mutex
+}
+
+// tunablesToC converts a RDMATunables value to the equivalent C struct.
+func tunablesToC(t RDMATunables) C.cuobj_rdma_tunables_t {
+	var ct C.cuobj_rdma_tunables_t
+	ct.num_dcis = C.int(t.NumDCIs)
+	ct.cq_depth = C.uint(t.CQDepth)
+	ct.dc_key = C.ulong(t.DCKey)
+	ct.service_level = C.int(t.ServiceLevel)
+	ct.timeout = C.uint8_t(t.Timeout)
+	ct.hop_limit = C.uint(t.HopLimit)
+	ct.pkey_index = C.int(t.PKeyIndex)
+	ct.delay_interval = C.uint32_t(t.DelayInterval)
+	ct.delay_mode = C.int(t.DelayMode)
+	ct.retry_cnt = C.uint8_t(t.RetryCount)
+	if t.QPResetOnFailure {
+		ct.qp_reset_on_failure = 1
+	}
+	ct.traffic_class = C.uint(t.TrafficClass)
+	ct.max_rd_atomic = C.int(t.MaxRdAtomic)
+	return ct
+}
+
+// NewServer creates a cuObjServer bound to the given RDMA IP and port.
+// Uses CUOBJ_PROTO_RDMA_DC_V1 (1001). If tunables is non-nil, the 4-argument
+// constructor is used so the tunable parameters apply to the initial session
+// started by the constructor. Pass nil to use library defaults.
+func NewServer(ip string, port uint16, tunables *RDMATunables) (*Server, error) {
+	cip := C.CString(ip)
+	defer C.free(unsafe.Pointer(cip))
+
+	var csrv *C.cuobj_server_t
+	if tunables != nil {
+		ct := tunablesToC(*tunables)
+		csrv = C.cuobj_server_create_with_config(cip, C.ushort(port), cuobjProtoRDMADCV1, &ct)
+	} else {
+		csrv = C.cuobj_server_create(cip, C.ushort(port), cuobjProtoRDMADCV1)
+	}
+	if csrv == nil {
+		return nil, fmt.Errorf("rdma: failed to create cuObjServer on %s:%d", ip, port)
+	}
+	srv := &Server{csrv: csrv}
+	// Some library versions start the session as part of construction;
+	// record that readiness so StartSession can be a no-op and Close/CloseSession
+	// use a consistent ownership model.
+	if srv.IsConnected() {
+		srv.sessionOpen = true
+	}
+	return srv, nil
+}
+
+// StartSession initiates the RDMA listening session.
+// StartSession must not be called concurrently with CloseSession or Close.
+func (s *Server) StartSession() error {
+	s.mu.Lock()
+	alreadyOpen := s.sessionOpen
+	s.mu.Unlock()
+	if alreadyOpen {
+		return nil
+	}
+	rc := C.cuobj_server_start_session(s.csrv)
+	if rc != 0 {
+		if s.IsConnected() {
+			s.mu.Lock()
+			s.sessionOpen = true
+			s.mu.Unlock()
+			return nil
+		}
+		return fmt.Errorf("rdma: startRDMASession failed (rc=%d)", rc)
+	}
+	s.mu.Lock()
+	s.sessionOpen = true
+	s.mu.Unlock()
+	return nil
+}
+
+// InitRDMAConfig applies RDMA tuning parameters. Must be called before StartSession.
+func (s *Server) InitRDMAConfig(t RDMATunables) error {
+	ct := tunablesToC(t)
+	if rc := C.cuobj_server_init_rdma_config(s.csrv, &ct); rc != 0 {
+		return fmt.Errorf("rdma: initRDMAConfigParams failed (rc=%d)", rc)
+	}
+	return nil
+}
+
+// IsConnected returns the RDMA connection status.
+func (s *Server) IsConnected() bool {
+	return C.cuobj_server_is_connected(s.csrv) != 0
+}
+
+// AllocHostBuffer allocates a 4KB-aligned host buffer of the given size.
+func (s *Server) AllocHostBuffer(size int) (unsafe.Pointer, error) {
+	if size <= 0 {
+		return nil, fmt.Errorf("rdma: allocHostBuffer size %d must be positive", size)
+	}
+	ptr := C.cuobj_server_alloc_host_buffer(s.csrv, C.size_t(size))
+	if ptr == nil {
+		return nil, fmt.Errorf("rdma: allocHostBuffer(%d) failed", size)
+	}
+	return ptr, nil
+}
+
+// FreeHostBuffer releases a buffer previously allocated by AllocHostBuffer.
+func (s *Server) FreeHostBuffer(ptr unsafe.Pointer) {
+	if ptr != nil {
+		C.cuobj_server_free_host_buffer(ptr)
+	}
+}
+
+// RegisterBuffer registers a host memory region for RDMA and returns a Buffer handle.
+func (s *Server) RegisterBuffer(ptr unsafe.Pointer, size int) (*Buffer, error) {
+	if ptr == nil {
+		return nil, errors.New("rdma: registerBuffer ptr must not be nil")
+	}
+	if size <= 0 {
+		return nil, fmt.Errorf("rdma: registerBuffer size %d must be positive", size)
+	}
+	cbuf := C.cuobj_server_register_buffer(s.csrv, ptr, C.size_t(size))
+	if cbuf == nil {
+		return nil, errors.New("rdma: registerBuffer failed")
+	}
+	return &Buffer{cbuf: cbuf, hostPtr: ptr, size: size}, nil
+}
+
+// DeregisterBuffer deregisters a previously registered RDMA buffer.
+func (s *Server) DeregisterBuffer(buf *Buffer) {
+	if buf != nil && buf.cbuf != nil {
+		C.cuobj_server_deregister_buffer(s.csrv, buf.cbuf)
+		buf.cbuf = nil
+		buf.hostPtr = nil
+		buf.size = 0
+	}
+}
+
+// AllocateChannel obtains a unique channel ID for concurrent RDMA operations.
+func (s *Server) AllocateChannel() (uint16, error) {
+	ch := C.cuobj_server_allocate_channel(s.csrv)
+	if ch == C.UINT16_MAX {
+		return 0, errors.New("rdma: no free channel IDs")
+	}
+	return uint16(ch), nil
+}
+
+// FreeChannel releases a previously allocated channel ID.
+func (s *Server) FreeChannel(id uint16) {
+	C.cuobj_server_free_channel(s.csrv, C.uint16_t(id))
+}
+
+// HandleGet performs an RDMA WRITE (server→client) to serve a GET request.
+// The local buffer must already contain the data to send.
+// Returns bytes transferred.
+func (s *Server) HandleGet(key string, buf *Buffer, remoteStart uint64, size int64, rdmaDescr string, channel uint16) (int64, error) {
+	if buf == nil || buf.cbuf == nil {
+		return 0, errors.New("rdma: invalid or deregistered buffer")
+	}
+	if size <= 0 {
+		return 0, fmt.Errorf("rdma: transfer size %d must be positive", size)
+	}
+	if size > MaxTransferSize {
+		return 0, fmt.Errorf("rdma: transfer size %d exceeds max %d", size, MaxTransferSize)
+	}
+
+	ckey := C.CString(key)
+	defer C.free(unsafe.Pointer(ckey))
+	cdescr := C.CString(rdmaDescr)
+	defer C.free(unsafe.Pointer(cdescr))
+
+	n := C.cuobj_server_handle_get(s.csrv, ckey, buf.cbuf,
+		C.uint64_t(remoteStart), C.size_t(size), cdescr, C.uint16_t(channel))
+	if n < 0 {
+		return 0, fmt.Errorf("rdma: handleGetObject failed (rc=%d)", n)
+	}
+	return int64(n), nil
+}
+
+// HandlePut performs an RDMA READ (client→server) to serve a PUT request.
+// After return, the local buffer contains the data read from the client.
+// Returns bytes transferred.
+func (s *Server) HandlePut(key string, buf *Buffer, remoteStart uint64, size int64, rdmaDescr string, channel uint16) (int64, error) {
+	if buf == nil || buf.cbuf == nil {
+		return 0, errors.New("rdma: invalid or deregistered buffer")
+	}
+	if size <= 0 {
+		return 0, fmt.Errorf("rdma: transfer size %d must be positive", size)
+	}
+	if size > MaxTransferSize {
+		return 0, fmt.Errorf("rdma: transfer size %d exceeds max %d", size, MaxTransferSize)
+	}
+
+	ckey := C.CString(key)
+	defer C.free(unsafe.Pointer(ckey))
+	cdescr := C.CString(rdmaDescr)
+	defer C.free(unsafe.Pointer(cdescr))
+
+	n := C.cuobj_server_handle_put(s.csrv, ckey, buf.cbuf,
+		C.uint64_t(remoteStart), C.size_t(size), cdescr, C.uint16_t(channel))
+	if n < 0 {
+		return 0, fmt.Errorf("rdma: handlePutObject failed (rc=%d)", n)
+	}
+	return int64(n), nil
+}
+
+// CloseSession tears down the RDMA session.
+func (s *Server) CloseSession() {
+	s.mu.Lock()
+	defer s.mu.Unlock()
+	if s.csrv != nil && s.sessionOpen {
+		C.cuobj_server_close_session(s.csrv)
+		s.sessionOpen = false
+	}
+}
+
+// Close destroys the cuObjServer instance. The Server must not be used afterward.
+func (s *Server) Close() {
+	s.mu.Lock()
+	defer s.mu.Unlock()
+	if s.csrv != nil {
+		if s.sessionOpen {
+			C.cuobj_server_close_session(s.csrv)
+			s.sessionOpen = false
+		}
+		C.cuobj_server_destroy(s.csrv)
+		s.csrv = nil
+	}
+}
@@ -0,0 +1,68 @@
+package rdma
+
+// MaxTransferSize is the cuObjServer limit per RDMA operation (1 GiB).
+const MaxTransferSize = 1 << 30
+
+// RDMATunables holds RDMA connection tuning parameters that map to
+// cuObjRDMATunableParam. Pass a non-nil pointer in CuServerOpts to apply
+// custom settings before the RDMA session starts; nil means library defaults.
+//
+// Use DefaultRDMATunables() to start from the library defaults and adjust
+// only the fields you care about.
+type RDMATunables struct {
+	// NumDCIs controls the maximum number of concurrent RDMA connections
+	// (Dynamic Connection Interfaces). Default: 128.
+	NumDCIs int
+	// CQDepth is the completion queue depth, limiting outstanding ops.
+	// Default: 640.
+	CQDepth uint32
+	// DCKey is the InfiniBand Dynamic Connection security key. All clients
+	// must use a matching key. Change this from the default in production.
+	// Default: 0xffeeddcc.
+	DCKey uint64
+	// ServiceLevel sets the IB QoS service level. Default: 0.
+	ServiceLevel int
+	// Timeout is the QP ACK timeout exponent: 4.096 * 2^Timeout µs.
+	// Default: 16 (~268 ms).
+	Timeout uint8
+	// HopLimit is the IB packet hop limit (analogous to IP TTL). Default: 4.
+	HopLimit uint32
+	// PKeyIndex is the IB partition key index. Default: 0.
+	PKeyIndex int
+	// DelayInterval is the polling delay in nanoseconds. Default: 5000.
+	DelayInterval uint32
+	// DelayMode selects the polling strategy:
+	//   0 = none, 1 = batch (default), 2 = per-entry, 3 = adaptive.
+	DelayMode int
+	// RetryCount is the QP retry count (0–7). Default: 7.
+	RetryCount uint8
+	// QPResetOnFailure controls whether the QP is reset after an RDMA failure.
+	// Default: true.
+	QPResetOnFailure bool
+	// TrafficClass sets the IB traffic class / DSCP bits. Default: 96.
+	TrafficClass uint32
+	// MaxRdAtomic is the max outstanding RDMA reads per DCI QP.
+	// 0 means auto-detect from device capabilities. Default: 0.
+	MaxRdAtomic int
+}
+
+// DefaultRDMATunables returns a RDMATunables pre-populated with the
+// cuObjServer library defaults. Use this as a starting point and override
+// only the fields you need.
+func DefaultRDMATunables() RDMATunables {
+	return RDMATunables{
+		NumDCIs:          128,
+		CQDepth:          640,
+		DCKey:            0xffeeddcc,
+		ServiceLevel:     0,
+		Timeout:          16,
+		HopLimit:         4,
+		PKeyIndex:        0,
+		DelayInterval:    5000,
+		DelayMode:        1, // CUOBJ_DELAY_BATCH
+		RetryCount:       7,
+		QPResetOnFailure: true,
+		TrafficClass:     96,
+		MaxRdAtomic:      0,
+	}
+}
@@ -0,0 +1,86 @@
+//go:build !(linux && amd64 && cgo)
+
+// Package rdma provides Go bindings to libcuobjserver.
+// This file is a stub for platforms without RDMA support.
+package rdma
+
+import (
+	"errors"
+	"unsafe"
+)
+
+var errNotSupported = errors.New("rdma: not supported on this platform")
+
+// Buffer wraps an RDMA-registered memory region (stub — always zero-valued).
+type Buffer struct {
+	hostPtr unsafe.Pointer
+	size    int
+}
+
+// HostPtr returns the underlying host memory pointer.
+func (b *Buffer) HostPtr() unsafe.Pointer { return b.hostPtr }
+
+// Size returns the buffer size in bytes.
+func (b *Buffer) Size() int { return b.size }
+
+// Slice returns the buffer contents as a Go byte slice.
+func (b *Buffer) Slice() []byte { return []byte{} }
+
+// Server wraps a cuObjServer instance (stub — always returns errNotSupported).
+type Server struct{}
+
+// ConfigureTelemetry is a no-op on platforms without RDMA support.
+func ConfigureTelemetry(debug bool) {}
+
+// DebugTelemetryEnabled always reports false on stub platforms.
+func DebugTelemetryEnabled() bool { return false }
+
+// NewServer always returns errNotSupported on this platform.
+func NewServer(ip string, port uint16, tunables *RDMATunables) (*Server, error) {
+	return nil, errNotSupported
+}
+
+// StartSession always returns errNotSupported on this platform.
+func (s *Server) StartSession() error { return errNotSupported }
+
+// InitRDMAConfig always returns errNotSupported on this platform.
+func (s *Server) InitRDMAConfig(t RDMATunables) error { return errNotSupported }
+
+// IsConnected always returns false on this platform.
+func (s *Server) IsConnected() bool { return false }
+
+// AllocHostBuffer always returns errNotSupported on this platform.
+func (s *Server) AllocHostBuffer(size int) (unsafe.Pointer, error) { return nil, errNotSupported }
+
+// FreeHostBuffer is a no-op on this platform.
+func (s *Server) FreeHostBuffer(ptr unsafe.Pointer) {}
+
+// RegisterBuffer always returns errNotSupported on this platform.
+func (s *Server) RegisterBuffer(ptr unsafe.Pointer, size int) (*Buffer, error) {
+	return nil, errNotSupported
+}
+
+// DeregisterBuffer is a no-op on this platform.
+func (s *Server) DeregisterBuffer(buf *Buffer) {}
+
+// AllocateChannel always returns errNotSupported on this platform.
+func (s *Server) AllocateChannel() (uint16, error) { return 0, errNotSupported }
+
+// FreeChannel is a no-op on this platform.
+func (s *Server) FreeChannel(id uint16) {}
+
+// HandleGet always returns errNotSupported on this platform.
+func (s *Server) HandleGet(key string, buf *Buffer, remoteStart uint64, size int64, rdmaDescr string, channel uint16) (int64, error) {
+	return 0, errNotSupported
+}
+
+// HandlePut always returns errNotSupported on this platform.
+func (s *Server) HandlePut(key string, buf *Buffer, remoteStart uint64, size int64, rdmaDescr string, channel uint16) (int64, error) {
+	return 0, errNotSupported
+}
+
+// CloseSession is a no-op on this platform.
+func (s *Server) CloseSession() {}
+
+// Close is a no-op on this platform.
+func (s *Server) Close() {}