feat/cuda: I4 - cuBLAS linear core (linear_f32 + linear_bf16) - #24
Open
magiodev wants to merge 11 commits into
Open
feat/cuda: I4 - cuBLAS linear core (linear_f32 + linear_bf16)#24magiodev wants to merge 11 commits into
magiodev wants to merge 11 commits into
Conversation
…l, 100 op stubs) - Add h3_cuda.cu: implements the existing plain-C h3_gpu.h API against CUDA/cuBLAS. h3_cuda_probe + h3_gpu_create/free/error real; all compute ops are stubs returning 'not yet implemented'. Metal backend (h3_gpu.m/h3_shaders.metal) preserved untouched. - Add h3_cuda.h (probe decl); h3.c probes via H3_CUDA guard. - Add h3_tokenizer.c C stub (Linux has no Foundation; Metal keeps h3_tokenizer.m). - h3_host.c: vImage high-quality scale guarded, portable bilinear fallback for CUDA build. - h3.c/h3_cli.c/h3_ffmpeg.c: Linux portability guards (st_mtimespec->st_mtim, arc4random_buf->getrandom, SSIZE_MAX define). - Makefile: cuda-spark/cuda-generic/cuda targets (nvcc + cuBLAS, .cuda.o objects, CUDA_ARCH=sm_121 for DGX Spark GB10), mirroring ds4's pattern.
…CK_MONOTONIC/strdup)
…urns cast to exact type
…removed prop.unifiedMemory)
…h in h3_cuda.cu nvcc compiles .cu as C++; without C linkage the GPU API symbols were mangled and the host C objects could not link. extern "C" guards are compile-time only and inert for the Metal build (.c/.m never define __cplusplus).
… command-buffer no-ops, stats - h3_gpu_tensor_new/from_*: cudaMalloc + host->device copy (f32/bf16/i8/u32) - h3_gpu_tensor_load_bf16/f32 + read_file_bf16/stream_file_bf16: chunked pread -> host staging -> device (1MB chunks; no host double-buffer of big weights) - h3_gpu_tensor_read/write_*: device<->host cudaMemcpy with bounds checks - h3_gpu_begin/continue/submit: no-ops (CUDA has no explicit command buffer) - h3_gpu_get_stats: allocated/live/peak bytes + tensor_allocations tracking - tensor_free: cudaFree + stats decrement; tensor struct carries owner - capability flags (is_m5/has_int8_mlp/has_nax_mlp) still return 0 (no int8/nax fast paths yet)
Real CUDA kernels (Metal formulas mirrored exactly): - activations: silu (f32/bf16), gelu (approx/exact, tanh/erf), geglu, swiglu (f32/bf16), silu_mul - casts: f32<->bf16 (round-to-nearest-even, matches Metal), clip - arithmetic: add/sub bf16, add_scaled f32 - norms: rms_norm/layer_norm (f32/bf16, block-per-row shared reduction), head_rms_norm, weight_norm - embedding bf16, scale_add f32, gate bf16 - copy f32/bf16 (device->device cudaMemcpy) BF16 helpers implemented to match Metal bit-for-bit. 50 compute ops remain stubs.
Row-major GEMM via cublasGemmEx(OP_T,OP_T): C=A@W^T. f32: CUDA_R_32F/CUBLAS_COMPUTE_32F. bf16: CUDA_R_16BF/CUBLAS_COMPUTE_32F (accum f32), bias added in f32, bf16 rounded once. h3_gpu_create now inits CUDA+cublas (handle stashed in dev_ctx), free destroys handle.
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
1 participant
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Stacked PR 4/4, on top of #20 (I1), #21 (I2), #23 (I3). Once those merge this diff narrows to I4 only.
I4: real matmul core via cuBLAS (cublasGemmEx row-major: C=A@W^T):
Build-gated on DGX Spark: make cuda-spark clean (binary 573KB).