jeffdaily commented on code in PR #1399: URL: https://github.com/apache/mahout/pull/1399#discussion_r3397736468
########## qdp/qdp-kernels/src/device.rs: ########## @@ -0,0 +1,418 @@ +// +// Licensed to the Apache Software Foundation (ASF) under one or more +// contributor license agreements. See the NOTICE file distributed with +// this work for additional information regarding copyright ownership. +// The ASF licenses this file to You 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. +// +// Copyright (c) 2026 Advanced Micro Devices, Inc. Review Comment: Done in 088041c -- per ASF policy, dropped the per-file AMD copyright and `Author:` lines from all 10 files (the ASF Apache-2.0 header is untouched) and added the AMD attribution to `NOTICE`. Git history preserves authorship. ########## qdp/qdp-core/src/gpu/cuda_ffi.rs: ########## @@ -37,65 +50,247 @@ pub(crate) struct CudaPointerAttributes { pub allocation_flags: u32, } -// CUDA error codes +// CUDA/HIP error codes (numerically identical for the codes used). pub(crate) const CUDA_SUCCESS: i32 = 0; -// Note: CUDA_ERROR_NOT_READY may be used in future optimizations for non-blocking event checks -// Reference: https://docs.nvidia.com/cuda/cuda-runtime-api/group__CUDART__TYPES.html#group__CUDART__TYPES_1g3f51e3575c2178246db0a94a430e0028 #[allow(dead_code)] pub(crate) const CUDA_ERROR_NOT_READY: i32 = 34; -unsafe extern "C" { - pub(crate) fn cudaHostAlloc(pHost: *mut *mut c_void, size: usize, flags: u32) -> i32; - pub(crate) fn cudaFreeHost(ptr: *mut c_void) -> i32; +// ---- CUDA backend: bind libcudart directly ---- +#[cfg(all(feature = "cuda", not(feature = "hip")))] +pub(crate) use cuda_rt::*; + +#[cfg(all(feature = "cuda", not(feature = "hip")))] +mod cuda_rt { + use super::CudaPointerAttributes; + use std::ffi::c_void; + + unsafe extern "C" { + pub(crate) fn cudaHostAlloc(pHost: *mut *mut c_void, size: usize, flags: u32) -> i32; + pub(crate) fn cudaFreeHost(ptr: *mut c_void) -> i32; + + #[allow(dead_code)] + pub(crate) fn cudaPointerGetAttributes( + attributes: *mut CudaPointerAttributes, + ptr: *const c_void, + ) -> i32; + + pub(crate) fn cudaMemGetInfo(free: *mut usize, total: *mut usize) -> i32; + + pub(crate) fn cudaMemcpyAsync( + dst: *mut c_void, + src: *const c_void, + count: usize, + kind: u32, + stream: *mut c_void, + ) -> i32; + + #[allow(dead_code)] + pub(crate) fn cudaMemcpy( + dst: *mut c_void, + src: *const c_void, + count: usize, + kind: u32, + ) -> i32; + + pub(crate) fn cudaEventCreateWithFlags(event: *mut *mut c_void, flags: u32) -> i32; + pub(crate) fn cudaEventRecord(event: *mut c_void, stream: *mut c_void) -> i32; + pub(crate) fn cudaEventDestroy(event: *mut c_void) -> i32; + pub(crate) fn cudaStreamWaitEvent( + stream: *mut c_void, + event: *mut c_void, + flags: u32, + ) -> i32; + pub(crate) fn cudaStreamSynchronize(stream: *mut c_void) -> i32; + + pub(crate) fn cudaMemsetAsync( + devPtr: *mut c_void, + value: i32, + count: usize, + stream: *mut c_void, + ) -> i32; + + #[allow(dead_code)] + pub(crate) fn cudaEventQuery(event: *mut c_void) -> i32; + pub(crate) fn cudaEventSynchronize(event: *mut c_void) -> i32; + pub(crate) fn cudaEventElapsedTime( + ms: *mut f32, + start: *mut c_void, + end: *mut c_void, + ) -> i32; + } +} + +// ---- HIP backend: bind libamdhip64, expose the same cuda* names ---- +#[cfg(feature = "hip")] +pub(crate) use hip_rt::*; + +// The wrapper functions deliberately keep the cuda* spelling so call sites are +// vendor-agnostic; suppress the snake_case lint for that intentional naming. +#[cfg(feature = "hip")] +#[allow(non_snake_case)] +mod hip_rt { + use super::{CUDA_MEMORY_TYPE_DEVICE, CUDA_MEMORY_TYPE_MANAGED, CudaPointerAttributes}; + use std::ffi::c_void; + + // hipMemoryType enum values are NOT guaranteed equal to CUDA's across ROCm + // releases (older HIP used Host=0/Device=1; the hip_runtime_api.h note flags + // this explicitly). So we read the real hipPointerAttribute_t and compare its + // `type` against the named hipMemoryType* constants rather than a magic + // number, then translate to the CUDA convention the caller expects. + const HIP_MEMORY_TYPE_DEVICE: i32 = 2; // hipMemoryTypeDevice Review Comment: Documented in dbee1e1 -- added a ROCm >= 6.0 prerequisite to DEVELOPMENT.md noting the `hipMemoryType` device-pointer check uses the ROCm 6+ convention (ROCm 5.x would reject valid device pointers). Tested on 7.2.1. ########## qdp/qdp-kernels/src/lib.rs: ########## @@ -20,6 +20,9 @@ use std::ffi::c_void; +pub mod device; +use device::{DeviceRepr, ValidAsZeroBits}; Review Comment: Added in dbee1e1 -- a `#[cfg(not(any(feature = "cuda", feature = "hip")))] compile_error!(...)` in device.rs, so that config now gives a clean diagnostic instead of the raw E0432. ########## qdp/DEVELOPMENT.md: ########## @@ -114,6 +114,43 @@ cd .. The first command is what `maturin develop --release` runs on CI; the second verifies tests type-check in the CUDA build. +### AMD GPU build (ROCm / HIP) Review Comment: Agreed -- the `gpu_rt` seam limits divergence, but without AMD runners the HIP path can still bit-rot. I'll put up a compile-only hipcc job (ROCm apt packages on an ubuntu runner) as a separate follow-up PR so the build is at least guarded. -- This is an automated message from the Apache Git Service. To respond to the message, please log on to GitHub and use the URL above to go to the specific comment. To unsubscribe, e-mail: [email protected] For queries about this service, please contact Infrastructure at: [email protected]
