diff --git a/cli/client.go b/cli/client.go index fd65e9b1..f534c814 100644 --- a/cli/client.go +++ b/cli/client.go @@ -222,6 +222,7 @@ func getClient(ctx *cli.Context, host string) (*minio.Client, error) { CustomMD5: md5simd.NewServer().NewHash, Transport: transport, TrailingHeaders: useTrailingHeaders.Load(), + EnableRDMA: ctx.String("rdma") != "", }) if err != nil { return nil, err diff --git a/cli/flags.go b/cli/flags.go index 6978ea3a..736a3c7b 100644 --- a/cli/flags.go +++ b/cli/flags.go @@ -181,6 +181,12 @@ var ioFlags = []cli.Flag{ Usage: "Use Kernel TLS (HTTPS) for transport if available", EnvVar: appNameUC + "_KTLS", }, + cli.StringFlag{ + Name: "rdma", + Usage: "Use S3-over-RDMA dispatch for PUT/GET. Values: \"cpu\" (host memory) or \"gpu\" (GPU-Direct, requires -tags=rdma + libcudart). Empty disables RDMA.", + EnvVar: appNameUC + "_RDMA", + Value: "", + }, cli.StringFlag{ Name: "region", Usage: "Specify a custom region", @@ -355,6 +361,7 @@ func getCommon(ctx *cli.Context, src func() generator.Source) bench.Common { Bucket: ctx.String("bucket"), Location: ctx.String("region"), PutOpts: putOpts, + RDMAMode: ctx.String("rdma"), DiscardOutput: noOps, ExtraOut: extra, RpsLimiter: rpsLimiter, diff --git a/go.mod b/go.mod index 94bc6995..6db1a007 100644 --- a/go.mod +++ b/go.mod @@ -2,7 +2,7 @@ module github.com/minio/warp go 1.25.0 -toolchain go1.26.3 +toolchain go1.26.4 tool ( github.com/golangci/golangci-lint/v2/cmd/golangci-lint @@ -24,12 +24,12 @@ require ( github.com/fatih/color v1.19.0 github.com/influxdata/influxdb-client-go/v2 v2.14.0 github.com/jfsmig/prng v0.0.2 - github.com/klauspost/compress v1.18.5 + github.com/klauspost/compress v1.18.6 github.com/minio/cli v1.24.2 github.com/minio/madmin-go/v4 v4.6.7 github.com/minio/mc v0.0.0-20251106162529-77f82e18b540 github.com/minio/md5-simd v1.1.2 - github.com/minio/minio-go/v7 v7.1.0 + github.com/minio/minio-go/v7 v7.2.0 github.com/minio/pkg/v3 v3.7.0 github.com/minio/websocket v1.6.0 github.com/muesli/termenv v0.16.0 @@ -167,7 +167,6 @@ require ( github.com/fzipp/gocyclo v0.6.0 // indirect github.com/ghostiam/protogetter v0.3.20 // indirect github.com/go-critic/go-critic v0.14.3 // indirect - github.com/go-ini/ini v1.67.0 // indirect github.com/go-jose/go-jose/v4 v4.1.4 // indirect github.com/go-logr/logr v1.4.3 // indirect github.com/go-logr/stdr v1.2.2 // indirect @@ -407,7 +406,7 @@ require ( google.golang.org/genproto/googleapis/rpc v0.0.0-20260128011058-8636f8732409 // indirect google.golang.org/grpc v1.79.3 // indirect google.golang.org/protobuf v1.36.11 // indirect - gopkg.in/ini.v1 v1.67.0 // indirect + gopkg.in/ini.v1 v1.67.2 // indirect gopkg.in/yaml.v2 v2.4.0 // indirect honnef.co/go/tools v0.7.0 // indirect mvdan.cc/gofumpt v0.9.2 // indirect diff --git a/go.sum b/go.sum index d55e23af..348e2072 100644 --- a/go.sum +++ b/go.sum @@ -394,8 +394,6 @@ github.com/ghostiam/protogetter v0.3.20 h1:oW7OPFit2FxZOpmMRPP9FffU4uUpfeE/rEdE1 github.com/ghostiam/protogetter v0.3.20/go.mod h1:FjIu5Yfs6FT391m+Fjp3fbAYJ6rkL/J6ySpZBfnODuI= github.com/go-critic/go-critic v0.14.3 h1:5R1qH2iFeo4I/RJU8vTezdqs08Egi4u5p6vOESA0pog= github.com/go-critic/go-critic v0.14.3/go.mod h1:xwntfW6SYAd7h1OqDzmN6hBX/JxsEKl5up/Y2bsxgVQ= -github.com/go-ini/ini v1.67.0 h1:z6ZrTEZqSWOTyH2FlglNbNgARyHG8oLW9gMELqKr06A= -github.com/go-ini/ini v1.67.0/go.mod h1:ByCAeIL28uOIIG0E3PJtZPDL8WnHpFKFOtgjp+3Ies8= github.com/go-jose/go-jose/v4 v4.1.4 h1:moDMcTHmvE6Groj34emNPLs/qtYXRVcd6S7NHbHz3kA= github.com/go-jose/go-jose/v4 v4.1.4/go.mod h1:x4oUasVrzR7071A4TnHLGSPpNOm2a21K9Kf04k1rs08= github.com/go-logr/logr v1.2.2/go.mod h1:jdQByPbusPIv2/zmleS9BjJVeZ6kBagPoEUsqbVz/1A= @@ -622,8 +620,8 @@ github.com/kkHAIKE/contextcheck v1.1.6 h1:7HIyRcnyzxL9Lz06NGhiKvenXq7Zw6Q0UQu/tt github.com/kkHAIKE/contextcheck v1.1.6/go.mod h1:3dDbMRNBFaq8HFXWC1JyvDSPm43CmE6IuHam8Wr0rkg= github.com/klauspost/asmfmt v1.3.2 h1:4Ri7ox3EwapiOjCki+hw14RyKk201CN4rzyCJRFLpK4= github.com/klauspost/asmfmt v1.3.2/go.mod h1:AG8TuvYojzulgDAMCnYn50l/5QV3Bs/tp6j0HLHbNSE= -github.com/klauspost/compress v1.18.5 h1:/h1gH5Ce+VWNLSWqPzOVn6XBO+vJbCNGvjoaGBFW2IE= -github.com/klauspost/compress v1.18.5/go.mod h1:cwPg85FWrGar70rWktvGQj8/hthj3wpl0PGDogxkrSQ= +github.com/klauspost/compress v1.18.6 h1:2jupLlAwFm95+YDR+NwD2MEfFO9d4z4Prjl1XXDjuao= +github.com/klauspost/compress v1.18.6/go.mod h1:cwPg85FWrGar70rWktvGQj8/hthj3wpl0PGDogxkrSQ= github.com/klauspost/cpuid/v2 v2.0.1/go.mod h1:FInQzS24/EEf25PyTYn52gqo7WaD8xa0213Md/qVLRg= github.com/klauspost/cpuid/v2 v2.0.9/go.mod h1:FInQzS24/EEf25PyTYn52gqo7WaD8xa0213Md/qVLRg= github.com/klauspost/cpuid/v2 v2.0.10/go.mod h1:g2LTdtYhdyuGPqyWyv7qRAmj1WBqxuObKfj5c0PQa7c= @@ -738,8 +736,8 @@ github.com/minio/mc v0.0.0-20251106162529-77f82e18b540 h1:OAeamQLGQyf7sT/JEocLpA github.com/minio/mc v0.0.0-20251106162529-77f82e18b540/go.mod h1:bqx15FhQpl5JfYU3yRM4iz2z2K6DiVSaPbj9P7trZZA= github.com/minio/md5-simd v1.1.2 h1:Gdi1DZK69+ZVMoNHRXJyNcxrMA4dSxoYHZSQbirFg34= github.com/minio/md5-simd v1.1.2/go.mod h1:MzdKDxYpY2BT9XQFocsiZf/NKVtR7nkE4RoEpN+20RM= -github.com/minio/minio-go/v7 v7.1.0 h1:QEt5IStDpxgGjEdtOgpiZ5QhmSl3ax7qy61vi2SwHO8= -github.com/minio/minio-go/v7 v7.1.0/go.mod h1:Dm7WS1AgLmBa0NcQD6SeJnJf+K/EUW3GR7Ks6olB3OA= +github.com/minio/minio-go/v7 v7.2.0 h1:RCJM0R1XOsRs+A3x3UCaf3ZYbByDaLjFeAi+YCQEPhs= +github.com/minio/minio-go/v7 v7.2.0/go.mod h1:EU9hENAStx/xXduNdrGO5e4X5vk19NtgB+RIPjZO8o0= github.com/minio/mux v1.9.0 h1:dWafQFyEfGhJvK6AwLOt83bIG5bxKxKJnKMCi0XAaoA= github.com/minio/mux v1.9.0/go.mod h1:1pAare17ZRL5GpmNL+9YmqHoWnLmMZF9C/ioUCfy0BQ= github.com/minio/pkg/v3 v3.7.0 h1:0aL3kyWUTwqKXMMq5SQG89UU6u4+Ov2kAdtRS9I8WSQ= @@ -963,6 +961,8 @@ github.com/ssgreg/nlreturn/v2 v2.2.1/go.mod h1:E/iiPB78hV7Szg2YfRgyIrk1AD6JVMTRk github.com/stbenjam/no-sprintf-host-port v0.3.1 h1:AyX7+dxI4IdLBPtDbsGAyqiTSLpCP9hWRrXQDU4Cm/g= github.com/stbenjam/no-sprintf-host-port v0.3.1/go.mod h1:ODbZesTCHMVKthBHskvUUexdcNHAQRXk9NpSsL8p/HQ= github.com/stretchr/objx v0.1.0/go.mod h1:HFkY916IF+rwdDfMAkV7OtwuqBVzrE8GR6GFx+wExME= +github.com/stretchr/objx v0.4.0/go.mod h1:YvHI0jy2hoMjB+UWwv71VJQ9isScKT/TqJzVSSt89Yw= +github.com/stretchr/objx v0.5.0/go.mod h1:Yh+to48EsGEfYuaHDzXPcE3xhTkx73EhmCGUpEOglKo= github.com/stretchr/objx v0.5.2 h1:xuMeJ0Sdp5ZMRXx/aWO6RZxdr3beISkG5/G/aIRr3pY= github.com/stretchr/objx v0.5.2/go.mod h1:FRsXN1f5AsAjCGJKqEizvkpNtU+EGNCLh3NxZ/8L+MA= github.com/stretchr/testify v1.2.2/go.mod h1:a8OnRcib4nhh0OaRAV+Yts87kKdq0PP7pXfy6kDkUVs= @@ -971,6 +971,8 @@ github.com/stretchr/testify v1.4.0/go.mod h1:j7eGeouHqKxXV5pUuKE4zz7dFj8WfuZ+81P github.com/stretchr/testify v1.6.1/go.mod h1:6Fq8oRcR53rry900zMqJjRRixrwX3KX962/h/Wwjteg= github.com/stretchr/testify v1.7.0/go.mod h1:6Fq8oRcR53rry900zMqJjRRixrwX3KX962/h/Wwjteg= github.com/stretchr/testify v1.7.1/go.mod h1:6Fq8oRcR53rry900zMqJjRRixrwX3KX962/h/Wwjteg= +github.com/stretchr/testify v1.8.0/go.mod h1:yNjHg4UonilssWZ8iaSj1OCr/vHnekPRkoO+kdMU+MU= +github.com/stretchr/testify v1.8.4/go.mod h1:sz/lmYIOXD/1dqDmKjjqLyZ2RngseejIcXlSw2iwfAo= github.com/stretchr/testify v1.11.1 h1:7s2iGBzp5EwR7/aIZr8ao5+dra3wiQyKjjFuvgVKu7U= github.com/stretchr/testify v1.11.1/go.mod h1:wZwfW3scLgRK+23gO65QZefKpKQRnfz6sD981Nm4B6U= github.com/subosito/gotenv v1.4.1 h1:jyEFiXpy21Wm81FBN71l9VoMMV8H8jG+qIK3GCpY6Qs= @@ -1291,8 +1293,8 @@ gopkg.in/evanphx/json-patch.v4 v4.12.0 h1:n6jtcsulIzXPJaxegRbvFNNrZDjbij7ny3gmSP gopkg.in/evanphx/json-patch.v4 v4.12.0/go.mod h1:p8EYWUEYMpynmqDbY58zCKCFZw8pRWMG4EsWvDvM72M= gopkg.in/inf.v0 v0.9.1 h1:73M5CoZyi3ZLMOyDlQh031Cx6N9NDJ2Vvfl76EDAgDc= gopkg.in/inf.v0 v0.9.1/go.mod h1:cWUDdTG/fYaXco+Dcufb5Vnc6Gp2YChqWtbxRZE0mXw= -gopkg.in/ini.v1 v1.67.0 h1:Dgnx+6+nfE+IfzjUEISNeydPJh9AXNNsWbGP9KzCsOA= -gopkg.in/ini.v1 v1.67.0/go.mod h1:pNLf8WUiyNEtQjuu5G5vTm06TEv9tsIgeAvK8hOrP4k= +gopkg.in/ini.v1 v1.67.2 h1:JtOSMb9OuaCZKr7h5D/h6iii14sK0hLbplTc6frx4Ss= +gopkg.in/ini.v1 v1.67.2/go.mod h1:x/cyOwCgZqOkJoDIJ3c1KNHMo10+nLGAhh+kn3Zizss= gopkg.in/urfave/cli.v1 v1.20.0/go.mod h1:vuBzUtMdQeixQj8LVd+/98pzhxNGQoyuPBlsXHOQNO0= gopkg.in/yaml.v2 v2.2.2/go.mod h1:hI93XBmqTisBFMUTm0b8Fm+jr3Dg1NNxqwp+5A1VGuI= gopkg.in/yaml.v2 v2.2.4/go.mod h1:hI93XBmqTisBFMUTm0b8Fm+jr3Dg1NNxqwp+5A1VGuI= diff --git a/pkg/bench/benchmark.go b/pkg/bench/benchmark.go index 8aa29892..74a603c6 100644 --- a/pkg/bench/benchmark.go +++ b/pkg/bench/benchmark.go @@ -53,6 +53,12 @@ type Common struct { // Default Put options. PutOpts minio.PutObjectOptions + // RDMAMode selects the per-op buffer source for minio-go's S3-over-RDMA + // dispatch. Empty disables RDMA. "cpu" allocates host memory. "gpu" + // allocates a CUDA device buffer (requires -tags=rdma,cuda + libcudart + // at build time) and lets the NIC GPU-Direct RDMA into it. + RDMAMode string + PrepareProgress chan float64 // Custom is returned to server if set by clients. diff --git a/pkg/bench/get.go b/pkg/bench/get.go index 4f38449e..512c3ff2 100644 --- a/pkg/bench/get.go +++ b/pkg/bench/get.go @@ -257,6 +257,46 @@ func (g *Get) Start(ctx context.Context, wait chan struct{}) error { if g.Versions > 1 { opts.VersionID = obj.VersionID } + if g.RDMAMode != RDMAModeOff { + // Allocate a host or GPU sink buffer (per --rdma) for + // minio-go's RDMA dispatch path; the server RDMA-writes + // the object directly into it. Object size is known + // from the corresponding upload (op.Size, already + // adjusted for random ranges above). + buf, berr := allocRDMABuf(g.RDMAMode, int(op.Size)) + if berr != nil { + g.Error("rdma alloc:", berr) + op.Err = berr.Error() + op.End = time.Now() + rcv <- op + cldone() + continue + } + opts.RDMABuffer = buf.ptr + opts.RDMABufferSize = buf.size + o, gerr := client.GetObject(nonTerm, g.Bucket, obj.Name, opts) + if gerr != nil { + g.Error("download error:", gerr) + op.Err = gerr.Error() + op.End = time.Now() + freeRDMABuf(buf) + rcv <- op + cldone() + continue + } + info, _ := o.Stat() + n := info.Size + op.End = time.Now() + if n != op.Size && op.Err == "" { + op.Err = fmt.Sprint("unexpected download size. want:", op.Size, ", got:", n) + g.Error(op.Err) + } + freeRDMABuf(buf) + rcv <- op + cldone() + o.Close() + continue + } o, err := client.GetObject(nonTerm, g.Bucket, obj.Name, opts) if err != nil { g.Error("download error:", err) diff --git a/pkg/bench/put.go b/pkg/bench/put.go index 93494a92..e84bb437 100644 --- a/pkg/bench/put.go +++ b/pkg/bench/put.go @@ -124,7 +124,27 @@ func (u *Put) Start(ctx context.Context, wait chan struct{}) error { var err error var res minio.UploadInfo if !u.PostObject { - res, err = client.PutObject(nonTerm, u.Bucket, obj.Name, obj.Reader, obj.Size, opts) + if u.RDMAMode != RDMAModeOff { + // Stage generator output into a CPU or GPU buffer + // (per --rdma) so minio-go's RDMA dispatch path can + // RDMA-WRITE it directly. Builds without -tags=rdma + // surface a clear error via minio-go's stub instead + // of silently falling back to HTTP. + buf, berr := allocRDMABuf(u.RDMAMode, int(obj.Size)) + if berr != nil { + err = berr + } else if serr := stageToRDMABuf(buf, obj.Reader); serr != nil { + err = fmt.Errorf("rdma upload prep: %w", serr) + freeRDMABuf(buf) + } else { + opts.RDMABuffer = buf.ptr + opts.RDMABufferSize = buf.size + res, err = client.PutObject(nonTerm, u.Bucket, obj.Name, nil, obj.Size, opts) + freeRDMABuf(buf) + } + } else { + res, err = client.PutObject(nonTerm, u.Bucket, obj.Name, obj.Reader, obj.Size, opts) + } } else { op.OpType = http.MethodPost var verID string diff --git a/pkg/bench/rdmabuf.go b/pkg/bench/rdmabuf.go new file mode 100644 index 00000000..a029e25d --- /dev/null +++ b/pkg/bench/rdmabuf.go @@ -0,0 +1,98 @@ +/* + * Warp (C) 2019-2026 MinIO, Inc. + * + * This program is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * (at your option) any later version. + * + * SPDX-License-Identifier: AGPL-3.0-or-later + */ + +package bench + +import ( + "errors" + "fmt" + "io" + "unsafe" +) + +// RDMAMode constants accepted by --rdma. +const ( + RDMAModeOff = "" + RDMAModeCPU = "cpu" + RDMAModeGPU = "gpu" +) + +// rdmaBuf is a per-op buffer registered with the NIC for S3-over-RDMA. +// For CPU mode, ptr points into pinned host memory (a Go []byte that we +// keep alive via the holder field). For GPU mode, ptr is a CUDA device +// pointer returned by cudaMalloc — see rdmabuf_rdma.go. +type rdmaBuf struct { + ptr unsafe.Pointer + size int + mode string + holder []byte // CPU mode: keeps the backing slice alive +} + +// allocRDMABuf returns a per-op buffer suitable for opts.RDMABuffer. +// `src`, when non-nil, is drained into the buffer (used by PUT to stage +// the generator output). For GPU mode the source data is uploaded via +// cudaMemcpy in stageToRDMABuf; for CPU mode it is a plain io.ReadFull. +func allocRDMABuf(mode string, size int) (*rdmaBuf, error) { + switch mode { + case RDMAModeCPU: + b := make([]byte, size) + if size == 0 { + return &rdmaBuf{mode: mode, size: 0, holder: b}, nil + } + return &rdmaBuf{ + ptr: unsafe.Pointer(&b[0]), + size: size, + mode: mode, + holder: b, + }, nil + case RDMAModeGPU: + return allocRDMAGPU(size) + default: + return nil, fmt.Errorf("rdma: unknown mode %q (want %q or %q)", + mode, RDMAModeCPU, RDMAModeGPU) + } +} + +// stageToRDMABuf drains src into the buffer. For CPU buffers this is a +// straight ReadFull into the backing slice. For GPU buffers the bytes +// are first read into a small CPU bounce buffer and then uploaded via +// cudaMemcpy host-to-device (see rdmabuf_rdma.go). +func stageToRDMABuf(b *rdmaBuf, src io.Reader) error { + if b == nil || b.size == 0 || src == nil { + return nil + } + switch b.mode { + case RDMAModeCPU: + _, err := io.ReadFull(src, b.holder) + return err + case RDMAModeGPU: + return stageToGPU(b, src) + default: + return fmt.Errorf("rdma: unknown mode %q", b.mode) + } +} + +// freeRDMABuf releases CUDA device memory; for CPU buffers it lets the +// Go runtime reclaim the backing slice once the rdmaBuf goes out of +// scope. +func freeRDMABuf(b *rdmaBuf) { + if b == nil { + return + } + if b.mode == RDMAModeGPU { + freeRDMAGPU(b) + } +} + +// errRDMAGPUUnsupported is returned by the stub GPU allocator built +// without -tags=rdma. +var errRDMAGPUUnsupported = errors.New( + "rdma=gpu requires building warp with -tags=rdma (libcudart + libminiocpp)") diff --git a/pkg/bench/rdmabuf_rdma.go b/pkg/bench/rdmabuf_rdma.go new file mode 100644 index 00000000..efea4d52 --- /dev/null +++ b/pkg/bench/rdmabuf_rdma.go @@ -0,0 +1,64 @@ +//go:build rdma + +/* + * Warp (C) 2019-2026 MinIO, Inc. + * + * SPDX-License-Identifier: AGPL-3.0-or-later + */ + +package bench + +// #cgo LDFLAGS: -lcudart +// #include +// #include +import "C" + +import ( + "fmt" + "io" + "unsafe" +) + +// allocRDMAGPU allocates a CUDA device buffer of size bytes. The +// returned rdmaBuf carries the device pointer in ptr; the NIC GPU- +// Direct RDMA-writes / reads into it via minio-go's RDMA dispatch. +func allocRDMAGPU(size int) (*rdmaBuf, error) { + if size <= 0 { + return &rdmaBuf{mode: RDMAModeGPU}, nil + } + var devPtr unsafe.Pointer + if rc := C.cudaMalloc(&devPtr, C.size_t(size)); rc != 0 { + return nil, fmt.Errorf("cudaMalloc(%d): cudaError=%d", size, int(rc)) + } + return &rdmaBuf{ptr: devPtr, size: size, mode: RDMAModeGPU}, nil +} + +// stageToGPU reads `size` bytes from src into a CPU bounce buffer, then +// cudaMemcpys host-to-device into the registered GPU buffer. The bounce +// buffer is reused per chunk to keep the per-op allocation budget low. +func stageToGPU(b *rdmaBuf, src io.Reader) error { + if b == nil || b.size == 0 || src == nil { + return nil + } + host := make([]byte, b.size) + if _, err := io.ReadFull(src, host); err != nil { + return err + } + rc := C.cudaMemcpy(b.ptr, + unsafe.Pointer(&host[0]), + C.size_t(b.size), + C.cudaMemcpyHostToDevice) + if rc != 0 { + return fmt.Errorf("cudaMemcpy H2D: cudaError=%d", int(rc)) + } + return nil +} + +// freeRDMAGPU releases the CUDA device buffer. +func freeRDMAGPU(b *rdmaBuf) { + if b == nil || b.ptr == nil { + return + } + C.cudaFree(b.ptr) + b.ptr = nil +} diff --git a/pkg/bench/rdmabuf_stub.go b/pkg/bench/rdmabuf_stub.go new file mode 100644 index 00000000..3ef46d21 --- /dev/null +++ b/pkg/bench/rdmabuf_stub.go @@ -0,0 +1,24 @@ +//go:build !rdma + +/* + * Warp (C) 2019-2026 MinIO, Inc. + * + * SPDX-License-Identifier: AGPL-3.0-or-later + */ + +package bench + +import "io" + +// allocRDMAGPU is unsupported on builds without -tags=rdma. Selecting +// --rdma=gpu produces a clear error rather than silently falling back +// to CPU memory. +func allocRDMAGPU(int) (*rdmaBuf, error) { + return nil, errRDMAGPUUnsupported +} + +func stageToGPU(*rdmaBuf, io.Reader) error { + return errRDMAGPUUnsupported +} + +func freeRDMAGPU(*rdmaBuf) {}