From 869caa6f3c329f77f744bd8ed00e8a0c9b1d6a6e Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 21:25:12 +0000 Subject: [PATCH 1/6] hipfile/test: Expose issue with temporary buffers and buffer offsets --- test/CMakeLists.txt | 1 + test/system/io.cpp | 69 +++++++++++++++++++++++++++++++++++++++++++++ 2 files changed, 70 insertions(+) create mode 100644 test/system/io.cpp diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt index cb605659..907e5fa3 100644 --- a/test/CMakeLists.txt +++ b/test/CMakeLists.txt @@ -22,6 +22,7 @@ set(SYSTEM_TEST_SOURCE_FILES system/buffer.cpp system/config.cpp system/driver.cpp + system/io.cpp system/main.cpp system/version.cpp ) diff --git a/test/system/io.cpp b/test/system/io.cpp new file mode 100644 index 00000000..b4931427 --- /dev/null +++ b/test/system/io.cpp @@ -0,0 +1,69 @@ +/* Copyright (c) Advanced Micro Devices, Inc. All rights reserved. + * + * SPDX-License-Identifier: MIT + */ + +#include "hipfile-warnings.h" +#include "hipfile.h" + +#include "test-common.h" +#include "test-options.h" + +#include +#include +#include + +extern SystemTestOptions test_env; + +HIPFILE_WARN_NO_GLOBAL_CTOR_OFF + +struct HipFileIo : public testing::Test { + + Tmpfile tmpfile; + size_t tmpfile_size; + hipFileHandle_t tmpfile_handle; + void *unregistered_device_buffer; + size_t unregistered_device_buffer_size; + + HipFileIo() + : tmpfile{test_env.ais_capable_dir}, tmpfile_size{1024 * 1024}, tmpfile_handle{nullptr}, + unregistered_device_buffer{nullptr}, unregistered_device_buffer_size{1024 * 1024} + { + } + + void SetUp() override + { + ASSERT_EQ(0, ftruncate(tmpfile.fd, static_cast(tmpfile_size))); + + hipFileDescr_t descr{}; + descr.type = hipFileHandleTypeOpaqueFD; + descr.handle.fd = tmpfile.fd; + + ASSERT_EQ(HIPFILE_SUCCESS, hipFileHandleRegister(&tmpfile_handle, &descr)); + + ASSERT_EQ(hipSuccess, hipMalloc(&unregistered_device_buffer, unregistered_device_buffer_size)); + } + + void TearDown() override + { + ASSERT_EQ(hipSuccess, hipFree(unregistered_device_buffer)); + + hipFileHandleDeregister(tmpfile_handle); + } +}; + +TEST_F(HipFileIo, ReadToUnregisteredBufferAtOffset) +{ + hoff_t io_buffer_offset{4096}; + size_t io_size{unregistered_device_buffer_size - static_cast(io_buffer_offset)}; + + // When we create a temporary buffer to handle the hipFileRead(), we pass in + // the io size to Buffer's constructor as the buffer's size. Before issuing + // the IO, the backends (fastpath/fallback) check if the IO would overflow + // the buffer. This check fails with temporary buffers because, + // size < size + offset when 0 < offset. + ASSERT_EQ(-hipFileInvalidValue, + hipFileRead(tmpfile_handle, unregistered_device_buffer, io_size, 0, io_buffer_offset)); +} + +HIPFILE_WARN_NO_GLOBAL_CTOR_ON From ff241ce202996632eaffaa228a28a58fc5b9ebfc Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 19:23:15 +0000 Subject: [PATCH 2/6] hipfile: Query buffer size when creating temporary buffers There are two options here. 1. hipfile does not store the size of the buffer in the Buffer object. This option relies on the bounds checking performed by the THUNK to ensure that the IO operation fits within the bounds of the buffer. This makes IO with an unregistered buffer a little faster because hipFile does not need to call hipMemGetAdressRange on every IO. However, if the THUNK detects that an IO would overflow the buffer the error is lost as the error propagates back up the stack and hipfile can only return -hipFileDriverError or -hipfileInternalError. 2. hipfile stores the size of the buffer in the Buffer object. hipFile performs its own bounds checking on each IO. This allows hipfile to return -hipfileInvalidValue when an IO would overflow the buffer. The downsides are IO to unregistered buffers now have to call hipMemGetAdressRange() on every IO and the bounds check duplicates the check done in the THUNK. hipfile will go with the second option because (1) IO with unregistered buffers is expected to e slower than with registered buffers and (2) it is better to give clients a more precise error (-hipfileInvalidValue). --- src/amd_detail/buffer.cpp | 24 +++++++++++++++++++++--- src/amd_detail/buffer.h | 10 +++++++++- src/amd_detail/state.h | 2 +- test/system/io.cpp | 13 ++++++++----- 4 files changed, 39 insertions(+), 10 deletions(-) diff --git a/src/amd_detail/buffer.cpp b/src/amd_detail/buffer.cpp index c01b43be..2f0a241a 100644 --- a/src/amd_detail/buffer.cpp +++ b/src/amd_detail/buffer.cpp @@ -60,6 +60,23 @@ Buffer::Buffer(const void *_buffer, size_t _length, int _flags, const PassKey &) : buffer{const_cast(_buffer)} +{ + if (!buffer) { + throw std::invalid_argument("Buffer pointer cannot be null."); + } + + hipPointerAttribute_t _attrs = Context::get()->hipPointerGetAttributes(buffer); + if (_attrs.type != hipMemoryTypeDevice) { + throw InvalidMemoryType(); + } + type = _attrs.type; + gpu_id = _attrs.device; + + HipMemAddressRange alp{Context::get()->hipMemGetAddressRange(buffer)}; + length = alp.size - (reinterpret_cast(alp.base) - reinterpret_cast(buffer)); +} + void * Buffer::getBuffer() const { @@ -130,12 +147,13 @@ BufferMap::getBuffer(const void *buf) shared_ptr BufferMap::getBuffer(const void *buf, size_t length, int flags) { + (void)flags; + auto itr = from_ptr.find(buf); if (from_ptr.end() == itr) { - // If the buffer hasn't been registered, use an unregistered - // temporary Buffer object - return std::shared_ptr(new Buffer(buf, length, flags, PassKey{})); + // Create a temporary buffer + return std::shared_ptr(new Buffer(buf, PassKey{})); } else { // If we found a registered buffer, it's an error if the diff --git a/src/amd_detail/buffer.h b/src/amd_detail/buffer.h index 67f377d7..298c627c 100644 --- a/src/amd_detail/buffer.h +++ b/src/amd_detail/buffer.h @@ -89,6 +89,14 @@ class Buffer : public IBuffer { /// @param k Key class instance (see passkey.h) Buffer(const void *buf, size_t length, int flags, const PassKey &k); + /// @brief Creates a buffer. + /// + /// The length of the buffer is determined by querying the hip runtime + /// + /// @param buf Buffer pointer + /// @param k Key class instance (see passkey.h) + Buffer(const void *buf, const PassKey &k); + private: /// @brief Pointer to a hip allocated buffer void *buffer; @@ -129,7 +137,7 @@ class BufferMap { virtual std::shared_ptr getBuffer(const void *buf); /// @brief Look up a registered buffer. Returns a temporary unregistered - /// buffer (of size length, using flags) if no matching buffer is found. + /// buffer if no matching buffer is found. /// @attention A shared_lock on HipFileMutex must be held /// @param buf Buffer pointer /// @param length Buffer length diff --git a/src/amd_detail/state.h b/src/amd_detail/state.h index d7d715b7..a819e964 100644 --- a/src/amd_detail/state.h +++ b/src/amd_detail/state.h @@ -102,7 +102,7 @@ class DriverState { virtual std::shared_ptr getBuffer(const void *buf); /// @brief Look up a registered buffer. Returns a temporary unregistered - /// buffer (of size length, using flags) if no matching buffer is found. + /// buffer if no matching buffer is found. /// @param [in] buf Buffer pointer /// @param [in] length Buffer length /// @param [in] flags Buffer flags (unused) diff --git a/test/system/io.cpp b/test/system/io.cpp index b4931427..e224e981 100644 --- a/test/system/io.cpp +++ b/test/system/io.cpp @@ -57,11 +57,14 @@ TEST_F(HipFileIo, ReadToUnregisteredBufferAtOffset) hoff_t io_buffer_offset{4096}; size_t io_size{unregistered_device_buffer_size - static_cast(io_buffer_offset)}; - // When we create a temporary buffer to handle the hipFileRead(), we pass in - // the io size to Buffer's constructor as the buffer's size. Before issuing - // the IO, the backends (fastpath/fallback) check if the IO would overflow - // the buffer. This check fails with temporary buffers because, - // size < size + offset when 0 < offset. + ASSERT_EQ(io_size, hipFileRead(tmpfile_handle, unregistered_device_buffer, io_size, 0, io_buffer_offset)); +} + +TEST_F(HipFileIo, ReadToUnregisteredBufferAtOffsetReturnsErrorIfOverflow) +{ + hoff_t io_buffer_offset{4096}; + size_t io_size{unregistered_device_buffer_size}; + ASSERT_EQ(-hipFileInvalidValue, hipFileRead(tmpfile_handle, unregistered_device_buffer, io_size, 0, io_buffer_offset)); } From abe25fd1e420bd2a77598a490e382dc26a39b18e Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 20:33:56 +0000 Subject: [PATCH 3/6] hipfile: Don't check buffer length on lookup The IO path checks that IO will not overflow. It is not clear what value this check provides. --- src/amd_detail/buffer.cpp | 6 +----- test/amd_detail/buffer.cpp | 10 ---------- test/amd_detail/hipfile-api.cpp | 7 ------- 3 files changed, 1 insertion(+), 22 deletions(-) diff --git a/src/amd_detail/buffer.cpp b/src/amd_detail/buffer.cpp index 2f0a241a..3d35c3f0 100644 --- a/src/amd_detail/buffer.cpp +++ b/src/amd_detail/buffer.cpp @@ -148,6 +148,7 @@ shared_ptr BufferMap::getBuffer(const void *buf, size_t length, int flags) { (void)flags; + (void)length; auto itr = from_ptr.find(buf); @@ -156,11 +157,6 @@ BufferMap::getBuffer(const void *buf, size_t length, int flags) return std::shared_ptr(new Buffer(buf, PassKey{})); } else { - // If we found a registered buffer, it's an error if the - // length parameter doesn't match what we found - if (itr->second->getLength() < length) { - throw std::invalid_argument("bad length parameter"); - } return itr->second; } } diff --git a/test/amd_detail/buffer.cpp b/test/amd_detail/buffer.cpp index 0eb34880..c3f58dcb 100644 --- a/test/amd_detail/buffer.cpp +++ b/test/amd_detail/buffer.cpp @@ -284,16 +284,6 @@ TEST_F(HipFileBuffer, get_buffer_returns_registered_buffer) Context::get()->getBuffer(nonnull_ptr)); } -TEST_F(HipFileBuffer, get_buffer_throws_if_length_larger_than_registered_length) -{ - StrictMock mhip; - size_t buffer_length = 0; - expect_buffer_registration(mhip, hipMemoryTypeDevice); - Context::get()->registerBuffer(nonnull_ptr, buffer_length, 0); - ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr, buffer_length + 1, 0), - std::invalid_argument); -} - TEST_F(HipFileBuffer, get_buffer_throws_on_getPointerAttributes_error) { StrictMock mhip; diff --git a/test/amd_detail/hipfile-api.cpp b/test/amd_detail/hipfile-api.cpp index dab50681..3ff32c09 100644 --- a/test/amd_detail/hipfile-api.cpp +++ b/test/amd_detail/hipfile-api.cpp @@ -174,13 +174,6 @@ TEST_P(HipFileIoParam, HipFileIoHandlesUnsupportedHipMemoryType) } } -TEST_P(HipFileIoParam, HipFileIoHandlesInvalidRegisteredBufferLength) -{ - StrictMock mhip; - ASSERT_EQ(hipFileIo(GetParam(), file_handle, bufptr, buflen + 1, 0, 0, mbackends), - -static_cast(hipFileInvalidValue)); -} - TEST_P(HipFileIoParam, HipFileIoHandlesInvalidFileHandle) { auto invalid_handle{reinterpret_cast(0xdeadbeef)}; From 9022b293f64aae1eae5a42ae0e9f78026e6b957d Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 21:01:12 +0000 Subject: [PATCH 4/6] hipfile: Rename DriverState::getBuffer() -> DriverState::getRegisteredBuffer() Rename getBuffer(ptr) to getRegisteredBuffer() now that BufferMap::getBuffer(ptr, length, flags) no longer uses length nor flags. This rename is necessary as a future commit will remove the unused parameters. --- src/amd_detail/state.cpp | 2 +- src/amd_detail/state.h | 2 +- test/amd_detail/buffer.cpp | 16 ++++++++-------- test/amd_detail/fallback.cpp | 6 +++--- test/amd_detail/mstate.h | 2 +- test/amd_detail/state_mt.cpp | 2 +- 6 files changed, 15 insertions(+), 15 deletions(-) diff --git a/src/amd_detail/state.cpp b/src/amd_detail/state.cpp index 85836a3f..d4f8c9c9 100644 --- a/src/amd_detail/state.cpp +++ b/src/amd_detail/state.cpp @@ -92,7 +92,7 @@ DriverState::deregisterBuffer(const void *buf) } shared_ptr -DriverState::getBuffer(const void *buf) +DriverState::getRegisteredBuffer(const void *buf) { unique_lock ulock{state_mutex}; diff --git a/src/amd_detail/state.h b/src/amd_detail/state.h index a819e964..e2709649 100644 --- a/src/amd_detail/state.h +++ b/src/amd_detail/state.h @@ -99,7 +99,7 @@ class DriverState { /// @brief Look up a registered buffer using the buffer pointer /// @param [in] buf Buffer pointer /// @return A registered buffer - virtual std::shared_ptr getBuffer(const void *buf); + virtual std::shared_ptr getRegisteredBuffer(const void *buf); /// @brief Look up a registered buffer. Returns a temporary unregistered /// buffer if no matching buffer is found. diff --git a/test/amd_detail/buffer.cpp b/test/amd_detail/buffer.cpp index c3f58dcb..bf8bbb42 100644 --- a/test/amd_detail/buffer.cpp +++ b/test/amd_detail/buffer.cpp @@ -210,7 +210,7 @@ TEST_F(HipFileBuffer, deregister_internal_get_prevents_deregister) expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(nonnull_ptr, 0, 0); { - auto buffer = Context::get()->getBuffer(nonnull_ptr); + auto buffer = Context::get()->getRegisteredBuffer(nonnull_ptr); ASSERT_THROW(Context::get()->deregisterBuffer(nonnull_ptr), BufferOperationsOutstanding); } Context::get()->deregisterBuffer(nonnull_ptr); @@ -222,7 +222,7 @@ TEST_F(HipFileBuffer, deregister_get_prevents_deregister) expect_buffer_registration(mhip, hipMemoryTypeDevice); ASSERT_EQ(hipFileBufRegister(nonnull_ptr, 0, 0), HIPFILE_SUCCESS); { - auto buffer = Context::get()->getBuffer(nonnull_ptr); + auto buffer = Context::get()->getRegisteredBuffer(nonnull_ptr); ASSERT_EQ(hipFileBufDeregister(nonnull_ptr), HipFileOpError(hipFileInternalError)); } ASSERT_EQ(hipFileBufDeregister(nonnull_ptr), HIPFILE_SUCCESS); @@ -230,7 +230,7 @@ TEST_F(HipFileBuffer, deregister_get_prevents_deregister) TEST_F(HipFileBuffer, get_not_registered) { - ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr), BufferNotRegistered); + ASSERT_THROW(Context::get()->getRegisteredBuffer(nonnull_ptr), BufferNotRegistered); } TEST_F(HipFileBuffer, get_internal_after_register) @@ -238,7 +238,7 @@ TEST_F(HipFileBuffer, get_internal_after_register) StrictMock mhip; expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(nonnull_ptr, 0, 0); - auto buffer = Context::get()->getBuffer(nonnull_ptr); + auto buffer = Context::get()->getRegisteredBuffer(nonnull_ptr); } TEST_F(HipFileBuffer, get_after_register) @@ -246,7 +246,7 @@ TEST_F(HipFileBuffer, get_after_register) StrictMock mhip; expect_buffer_registration(mhip, hipMemoryTypeDevice); ASSERT_EQ(hipFileBufRegister(nonnull_ptr, 0, 0), HIPFILE_SUCCESS); - auto buffer = Context::get()->getBuffer(nonnull_ptr); + auto buffer = Context::get()->getRegisteredBuffer(nonnull_ptr); } TEST_F(HipFileBuffer, get_internal_after_deregister) @@ -255,7 +255,7 @@ TEST_F(HipFileBuffer, get_internal_after_deregister) expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(nonnull_ptr, 0, 0); Context::get()->deregisterBuffer(nonnull_ptr); - ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr), BufferNotRegistered); + ASSERT_THROW(Context::get()->getRegisteredBuffer(nonnull_ptr), BufferNotRegistered); } TEST_F(HipFileBuffer, get_after_deregister) @@ -264,7 +264,7 @@ TEST_F(HipFileBuffer, get_after_deregister) expect_buffer_registration(mhip, hipMemoryTypeDevice); ASSERT_EQ(hipFileBufRegister(nonnull_ptr, 0, 0), HIPFILE_SUCCESS); ASSERT_EQ(hipFileBufDeregister(nonnull_ptr), HIPFILE_SUCCESS); - ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr), BufferNotRegistered); + ASSERT_THROW(Context::get()->getRegisteredBuffer(nonnull_ptr), BufferNotRegistered); } TEST_F(HipFileBuffer, get_buffer_makes_temporary_buffer) @@ -281,7 +281,7 @@ TEST_F(HipFileBuffer, get_buffer_returns_registered_buffer) expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(nonnull_ptr, 0, 0); ASSERT_EQ(Context::get()->getBuffer(nonnull_ptr, 0, 0), - Context::get()->getBuffer(nonnull_ptr)); + Context::get()->getRegisteredBuffer(nonnull_ptr)); } TEST_F(HipFileBuffer, get_buffer_throws_on_getPointerAttributes_error) diff --git a/test/amd_detail/fallback.cpp b/test/amd_detail/fallback.cpp index 9060f85c..dffa530c 100644 --- a/test/amd_detail/fallback.cpp +++ b/test/amd_detail/fallback.cpp @@ -119,7 +119,7 @@ struct FallbackIo : public HipFileOpened { expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(buffer_data.data(), buffer_data.size(), 0); - buffer = Context::get()->getBuffer(buffer_data.data()); + buffer = Context::get()->getRegisteredBuffer(buffer_data.data()); expect_file_registration(msys, mlibmounthelper); file = Context::get()->getFile(Context::get()->registerFile(0xBADF00D)); @@ -182,7 +182,7 @@ struct FallbackParam : ::testing::TestWithParam { expect_buffer_registration(mhip, hipMemoryTypeDevice); void *buf = reinterpret_cast(0xFEFEFEFE); Context::get()->registerBuffer(buf, 4096, 0); - buffer = Context::get()->getBuffer(buf); + buffer = Context::get()->getRegisteredBuffer(buf); expect_file_registration(msys, mlibmounthelper); file = Context::get()->getFile(Context::get()->registerFile(0xBADF00D)); @@ -248,7 +248,7 @@ TEST_P(FallbackParam, fallback_io_truncates_size_to_MAX_RW_COUNT) expect_buffer_registration(mhip, hipMemoryTypeDevice); auto buf = reinterpret_cast(0xABABABAB); Context::get()->registerBuffer(buf, MAX_RW_COUNT + 1, 0); - auto big_buffer = Context::get()->getBuffer(buf); + auto big_buffer = Context::get()->getRegisteredBuffer(buf); EXPECT_CALL(msys, mmap).WillOnce(testing::Return(reinterpret_cast(0xFEFEFEFE))); switch (io_type) { diff --git a/test/amd_detail/mstate.h b/test/amd_detail/mstate.h index d3f618f0..94fc3e60 100644 --- a/test/amd_detail/mstate.h +++ b/test/amd_detail/mstate.h @@ -32,7 +32,7 @@ class MDriverState : public DriverState { MOCK_METHOD(std::shared_ptr, getBatchContext, (hipFileBatchHandle_t handle), (override)); MOCK_METHOD(void, registerBuffer, (const void *buf, size_t length, int flags), (override)); MOCK_METHOD(void, deregisterBuffer, (const void *buf), (override)); - MOCK_METHOD(std::shared_ptr, getBuffer, (const void *buf), (override)); + MOCK_METHOD(std::shared_ptr, getRegisteredBuffer, (const void *buf), (override)); MOCK_METHOD(std::shared_ptr, getBuffer, (const void *buf, size_t length, int flags), (override)); MOCK_METHOD(hipFileHandle_t, registerFile, (UnregisteredFile && uf), (override)); MOCK_METHOD(void, deregisterFile, (hipFileHandle_t fh), (override)); diff --git a/test/amd_detail/state_mt.cpp b/test/amd_detail/state_mt.cpp index d814fa19..14db95b3 100644 --- a/test/amd_detail/state_mt.cpp +++ b/test/amd_detail/state_mt.cpp @@ -243,7 +243,7 @@ thread_function(int id) uniform_int_distribution vec_dist{0, buffers.size() - 1}; size_t idx = vec_dist(gen); - auto data = ds->getBuffer(buffers[idx]); + auto data = ds->getRegisteredBuffer(buffers[idx]); // TODO: Maintain a data collection to ensure there are no races // on the data From 3014294abbb01c8f661a4b9a3b6aa68bcd8c7756 Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 21:01:12 +0000 Subject: [PATCH 5/6] hipfile: Rename BufferMap::getBuffer() -> BufferMap::getRegisteredBuffer() Rename getBuffer(ptr) to getRegisteredBuffer() now that BufferMap::getBuffer(ptr, length, flags) no longer uses length nor flags. This rename is necessary as a future commit will remove the unused parameters. --- src/amd_detail/buffer.cpp | 2 +- src/amd_detail/buffer.h | 2 +- src/amd_detail/state.cpp | 2 +- test/amd_detail/mbuffer.h | 2 +- 4 files changed, 4 insertions(+), 4 deletions(-) diff --git a/src/amd_detail/buffer.cpp b/src/amd_detail/buffer.cpp index 3d35c3f0..dec1427c 100644 --- a/src/amd_detail/buffer.cpp +++ b/src/amd_detail/buffer.cpp @@ -134,7 +134,7 @@ BufferMap::deregisterBuffer(const void *buf) } shared_ptr -BufferMap::getBuffer(const void *buf) +BufferMap::getRegisteredBuffer(const void *buf) { auto itr = from_ptr.find(buf); if (from_ptr.end() == itr) { diff --git a/src/amd_detail/buffer.h b/src/amd_detail/buffer.h index 298c627c..954b8408 100644 --- a/src/amd_detail/buffer.h +++ b/src/amd_detail/buffer.h @@ -134,7 +134,7 @@ class BufferMap { /// @attention A shared_lock on HipFileMutex must be held /// @param buf Buffer pointer /// @return A registered buffer - virtual std::shared_ptr getBuffer(const void *buf); + virtual std::shared_ptr getRegisteredBuffer(const void *buf); /// @brief Look up a registered buffer. Returns a temporary unregistered /// buffer if no matching buffer is found. diff --git a/src/amd_detail/state.cpp b/src/amd_detail/state.cpp index d4f8c9c9..dd6b1ad4 100644 --- a/src/amd_detail/state.cpp +++ b/src/amd_detail/state.cpp @@ -100,7 +100,7 @@ DriverState::getRegisteredBuffer(const void *buf) throw DriverNotInitialized(); } - return buffer_map->getBuffer(buf); + return buffer_map->getRegisteredBuffer(buf); } shared_ptr diff --git a/test/amd_detail/mbuffer.h b/test/amd_detail/mbuffer.h index 9c1eb206..fdeffb9c 100644 --- a/test/amd_detail/mbuffer.h +++ b/test/amd_detail/mbuffer.h @@ -31,7 +31,7 @@ class MBufferMap : public BufferMap { } MOCK_METHOD(void, registerBuffer, (const void *bufptr, size_t length, int flags), (override)); MOCK_METHOD(void, deregisterBuffer, (const void *bufptr), (override)); - MOCK_METHOD(std::shared_ptr, getBuffer, (const void *bufptr), (override)); + MOCK_METHOD(std::shared_ptr, getRegisteredBuffer, (const void *bufptr), (override)); MOCK_METHOD(std::shared_ptr, getBuffer, (const void *bufptr, size_t length, int flags), (override)); MOCK_METHOD(void, clear, (), (override)); From 1593911c5048441efad27979afcaf4c929b779e5 Mon Sep 17 00:00:00 2001 From: Kurt McMillan Date: Fri, 30 Jan 2026 21:23:20 +0000 Subject: [PATCH 6/6] hipfile: Remove unused length and flags parameters from buffer lookup functions --- src/amd_detail/batch/batch.cpp | 4 ++-- src/amd_detail/buffer.cpp | 5 +---- src/amd_detail/buffer.h | 5 ++--- src/amd_detail/hipfile.cpp | 2 +- src/amd_detail/state.cpp | 14 ++++++-------- src/amd_detail/state.h | 12 ++++-------- test/amd_detail/buffer.cpp | 6 +++--- test/amd_detail/hipfile-api.cpp | 9 +++------ test/amd_detail/mbuffer.h | 3 +-- test/amd_detail/mstate.h | 5 ++--- test/amd_detail/state_mt.cpp | 7 +++---- 11 files changed, 28 insertions(+), 44 deletions(-) diff --git a/src/amd_detail/batch/batch.cpp b/src/amd_detail/batch/batch.cpp index 7ff7c14f..1c228735 100644 --- a/src/amd_detail/batch/batch.cpp +++ b/src/amd_detail/batch/batch.cpp @@ -122,8 +122,8 @@ BatchContext::submit_operations(const hipFileIOParams_t *params, unsigned num_pa auto param_copy = std::make_unique(params[i]); // flags currently unused. Ambiguous if flags in hipFileBatchIOSubmit is for buffer or // file flags. - auto [_file, _buffer] = Context::get()->getFileAndBuffer( - param_copy->fh, param_copy->u.batch.devPtr_base, param_copy->u.batch.size, 0); + auto [_file, _buffer] = + Context::get()->getFileAndBuffer(param_copy->fh, param_copy->u.batch.devPtr_base); auto op = std::make_shared(std::move(param_copy), _buffer, _file); pending_ops.push_back(op); diff --git a/src/amd_detail/buffer.cpp b/src/amd_detail/buffer.cpp index dec1427c..5c0e768b 100644 --- a/src/amd_detail/buffer.cpp +++ b/src/amd_detail/buffer.cpp @@ -145,11 +145,8 @@ BufferMap::getRegisteredBuffer(const void *buf) } shared_ptr -BufferMap::getBuffer(const void *buf, size_t length, int flags) +BufferMap::getBuffer(const void *buf) { - (void)flags; - (void)length; - auto itr = from_ptr.find(buf); if (from_ptr.end() == itr) { diff --git a/src/amd_detail/buffer.h b/src/amd_detail/buffer.h index 954b8408..bcde5733 100644 --- a/src/amd_detail/buffer.h +++ b/src/amd_detail/buffer.h @@ -137,12 +137,11 @@ class BufferMap { virtual std::shared_ptr getRegisteredBuffer(const void *buf); /// @brief Look up a registered buffer. Returns a temporary unregistered - /// buffer if no matching buffer is found. + /// buffer if no registered buffer is found. /// @attention A shared_lock on HipFileMutex must be held /// @param buf Buffer pointer - /// @param length Buffer length /// @return A registered or temporary unregistered buffer - virtual std::shared_ptr getBuffer(const void *buf, size_t length, int flags); + virtual std::shared_ptr getBuffer(const void *buf); virtual void clear(); diff --git a/src/amd_detail/hipfile.cpp b/src/amd_detail/hipfile.cpp index 7c683c57..fad92290 100644 --- a/src/amd_detail/hipfile.cpp +++ b/src/amd_detail/hipfile.cpp @@ -158,7 +158,7 @@ ssize_t hipFileIo(IoType type, hipFileHandle_t fh, const void *buffer_base, size_t size, hoff_t file_offset, hoff_t buffer_offset, const vector> &backends) try { - auto [file, buffer] = Context::get()->getFileAndBuffer(fh, buffer_base, size, 0); + auto [file, buffer] = Context::get()->getFileAndBuffer(fh, buffer_base); int score{-1}; std::shared_ptr backend{}; diff --git a/src/amd_detail/state.cpp b/src/amd_detail/state.cpp index dd6b1ad4..197b795e 100644 --- a/src/amd_detail/state.cpp +++ b/src/amd_detail/state.cpp @@ -104,13 +104,13 @@ DriverState::getRegisteredBuffer(const void *buf) } shared_ptr -DriverState::getBuffer(const void *buf, size_t length, int flags) +DriverState::getBuffer(const void *buf) { // NOTE: This mutex only protects the map, so we'll // also need to protect the data shared_lock slock{state_mutex}; - return buffer_map->getBuffer(buf, length, flags); + return buffer_map->getBuffer(buf); } // @@ -190,7 +190,7 @@ DriverState::getStream(hipStream_t hip_stream) // file_buffer_pair -DriverState::getFileAndBuffer(hipFileHandle_t fh, const void *buf, size_t length, int flags) +DriverState::getFileAndBuffer(hipFileHandle_t fh, const void *buf) { // NOTE: This mutex only protects the map, so we'll // also need to protect the data @@ -200,7 +200,7 @@ DriverState::getFileAndBuffer(hipFileHandle_t fh, const void *buf, size_t length throw DriverNotInitialized(); } - return {file_map->getFile(fh), buffer_map->getBuffer(buf, length, flags)}; + return {file_map->getFile(fh), buffer_map->getBuffer(buf)}; } // @@ -208,8 +208,7 @@ DriverState::getFileAndBuffer(hipFileHandle_t fh, const void *buf, size_t length // file_buffer_stream_tuple -DriverState::getFileBufferAndStream(hipFileHandle_t fh, const void *buf, size_t length, int flags, - hipStream_t hipStream) +DriverState::getFileBufferAndStream(hipFileHandle_t fh, const void *buf, hipStream_t hipStream) { shared_lock slock{state_mutex}; @@ -217,8 +216,7 @@ DriverState::getFileBufferAndStream(hipFileHandle_t fh, const void *buf, size_t throw DriverNotInitialized(); } - return {file_map->getFile(fh), buffer_map->getBuffer(buf, length, flags), - stream_map->getStream(hipStream)}; + return {file_map->getFile(fh), buffer_map->getBuffer(buf), stream_map->getStream(hipStream)}; } // diff --git a/src/amd_detail/state.h b/src/amd_detail/state.h index e2709649..99200f8e 100644 --- a/src/amd_detail/state.h +++ b/src/amd_detail/state.h @@ -104,10 +104,8 @@ class DriverState { /// @brief Look up a registered buffer. Returns a temporary unregistered /// buffer if no matching buffer is found. /// @param [in] buf Buffer pointer - /// @param [in] length Buffer length - /// @param [in] flags Buffer flags (unused) /// @return A registered or temporary unregistered buffer - virtual std::shared_ptr getBuffer(const void *buf, size_t length, int flags); + virtual std::shared_ptr getBuffer(const void *buf); // // File interface @@ -158,19 +156,17 @@ class DriverState { /// This combined file + buffer getter reduces the number of lock calls. /// /// Like the buffer getter, this function emits a temporary unregistered buffer - /// (of size length, using flags) if no matching buffer is found. + /// if no matching registered buffer is found. /// /// @param [in] fh File handle /// @param [in] buf Buffer pointer - /// @param [in] length Buffer length - /// @param [in] flags Buffer flags (unused) - virtual file_buffer_pair getFileAndBuffer(hipFileHandle_t fh, const void *buf, size_t length, int flags); + virtual file_buffer_pair getFileAndBuffer(hipFileHandle_t fh, const void *buf); // // Buffer, file, and stream calls // virtual file_buffer_stream_tuple getFileBufferAndStream(hipFileHandle_t fh, const void *buf, - size_t length, int flags, hipStream_t hipStream); + hipStream_t hipStream); // // Reference counts diff --git a/test/amd_detail/buffer.cpp b/test/amd_detail/buffer.cpp index bf8bbb42..91ac41f2 100644 --- a/test/amd_detail/buffer.cpp +++ b/test/amd_detail/buffer.cpp @@ -271,7 +271,7 @@ TEST_F(HipFileBuffer, get_buffer_makes_temporary_buffer) { StrictMock mhip; expect_buffer_registration(mhip, hipMemoryTypeDevice); - auto buffer = Context::get()->getBuffer(nonnull_ptr, 0, 0); + auto buffer = Context::get()->getBuffer(nonnull_ptr); ASSERT_EQ(buffer.use_count(), 1); } @@ -280,7 +280,7 @@ TEST_F(HipFileBuffer, get_buffer_returns_registered_buffer) StrictMock mhip; expect_buffer_registration(mhip, hipMemoryTypeDevice); Context::get()->registerBuffer(nonnull_ptr, 0, 0); - ASSERT_EQ(Context::get()->getBuffer(nonnull_ptr, 0, 0), + ASSERT_EQ(Context::get()->getBuffer(nonnull_ptr), Context::get()->getRegisteredBuffer(nonnull_ptr)); } @@ -288,7 +288,7 @@ TEST_F(HipFileBuffer, get_buffer_throws_on_getPointerAttributes_error) { StrictMock mhip; EXPECT_CALL(mhip, hipPointerGetAttributes).WillOnce(testing::Throw(Hip::RuntimeError(hipErrorUnknown))); - ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr, 1, 0), Hip::RuntimeError); + ASSERT_THROW(Context::get()->getBuffer(nonnull_ptr), Hip::RuntimeError); } HIPFILE_WARN_NO_GLOBAL_CTOR_ON diff --git a/test/amd_detail/hipfile-api.cpp b/test/amd_detail/hipfile-api.cpp index 3ff32c09..73367ffe 100644 --- a/test/amd_detail/hipfile-api.cpp +++ b/test/amd_detail/hipfile-api.cpp @@ -235,8 +235,7 @@ TEST_P(HipFileIoBackendSelectionParam, HipFileIoThrowsIfThereAreNoBackends) { auto backends{std::vector>()}; - EXPECT_CALL(mds, getFileAndBuffer(handle, buffer, io_size, flags)) - .WillOnce(Return(file_buffer_pair{mfile, mbuffer})); + EXPECT_CALL(mds, getFileAndBuffer(handle, buffer)).WillOnce(Return(file_buffer_pair{mfile, mbuffer})); EXPECT_CALL(mds, getBackends).WillOnce(Return(backends)); switch (io_type) { @@ -257,8 +256,7 @@ TEST_P(HipFileIoBackendSelectionParam, HipFileIoThrowsIfAllBackendsRejectTheIO) { std::vector> backends{mbe1, mbe2, mbe3}; - EXPECT_CALL(mds, getFileAndBuffer(handle, buffer, io_size, flags)) - .WillOnce(Return(file_buffer_pair{mfile, mbuffer})); + EXPECT_CALL(mds, getFileAndBuffer(handle, buffer)).WillOnce(Return(file_buffer_pair{mfile, mbuffer})); EXPECT_CALL(mds, getBackends).WillOnce(Return(backends)); EXPECT_CALL(*mbe1, score(Eq(mfile), Eq(mbuffer), io_size, file_offset, buffer_offset)) .WillOnce(Return(-1)); @@ -285,8 +283,7 @@ TEST_P(HipFileIoBackendSelectionParam, HipFileIoIssuesIoToHighestScoringBackend) { std::vector> backends{mbe1, mbe2, mbe3}; - EXPECT_CALL(mds, getFileAndBuffer(handle, buffer, io_size, flags)) - .WillOnce(Return(file_buffer_pair{mfile, mbuffer})); + EXPECT_CALL(mds, getFileAndBuffer(handle, buffer)).WillOnce(Return(file_buffer_pair{mfile, mbuffer})); EXPECT_CALL(mds, getBackends).WillOnce(Return(backends)); EXPECT_CALL(*mbe1, score(Eq(mfile), Eq(mbuffer), io_size, file_offset, buffer_offset)) .WillOnce(Return(0)); diff --git a/test/amd_detail/mbuffer.h b/test/amd_detail/mbuffer.h index fdeffb9c..c9840f36 100644 --- a/test/amd_detail/mbuffer.h +++ b/test/amd_detail/mbuffer.h @@ -32,8 +32,7 @@ class MBufferMap : public BufferMap { MOCK_METHOD(void, registerBuffer, (const void *bufptr, size_t length, int flags), (override)); MOCK_METHOD(void, deregisterBuffer, (const void *bufptr), (override)); MOCK_METHOD(std::shared_ptr, getRegisteredBuffer, (const void *bufptr), (override)); - MOCK_METHOD(std::shared_ptr, getBuffer, (const void *bufptr, size_t length, int flags), - (override)); + MOCK_METHOD(std::shared_ptr, getBuffer, (const void *bufptr), (override)); MOCK_METHOD(void, clear, (), (override)); }; diff --git a/test/amd_detail/mstate.h b/test/amd_detail/mstate.h index 94fc3e60..2d281954 100644 --- a/test/amd_detail/mstate.h +++ b/test/amd_detail/mstate.h @@ -33,12 +33,11 @@ class MDriverState : public DriverState { MOCK_METHOD(void, registerBuffer, (const void *buf, size_t length, int flags), (override)); MOCK_METHOD(void, deregisterBuffer, (const void *buf), (override)); MOCK_METHOD(std::shared_ptr, getRegisteredBuffer, (const void *buf), (override)); - MOCK_METHOD(std::shared_ptr, getBuffer, (const void *buf, size_t length, int flags), (override)); + MOCK_METHOD(std::shared_ptr, getBuffer, (const void *buf), (override)); MOCK_METHOD(hipFileHandle_t, registerFile, (UnregisteredFile && uf), (override)); MOCK_METHOD(void, deregisterFile, (hipFileHandle_t fh), (override)); MOCK_METHOD(std::shared_ptr, getFile, (hipFileHandle_t fh), (override)); - MOCK_METHOD(file_buffer_pair, getFileAndBuffer, - (hipFileHandle_t fh, const void *buf, size_t length, int flags), (override)); + MOCK_METHOD(file_buffer_pair, getFileAndBuffer, (hipFileHandle_t fh, const void *buf), (override)); MOCK_METHOD(void, incrRefCount, (), (override)); MOCK_METHOD(void, decrRefCount, (), (override)); MOCK_METHOD(int64_t, getRefCount, (), (override, const)); diff --git a/test/amd_detail/state_mt.cpp b/test/amd_detail/state_mt.cpp index 14db95b3..79974108 100644 --- a/test/amd_detail/state_mt.cpp +++ b/test/amd_detail/state_mt.cpp @@ -257,7 +257,7 @@ thread_function(int id) uniform_int_distribution vec_dist{0, buffers.size() - 1}; size_t idx = vec_dist(gen); - auto data = ds->getBuffer(buffers[idx], ALLOC_SIZE, 0); + auto data = ds->getBuffer(buffers[idx]); // TODO: Maintain a data collection to ensure there are no races // on the data @@ -291,7 +291,7 @@ thread_function(int id) idx = buf_dist(gen); auto buf = buffers[idx]; - auto [file, buffer] = ds->getFileAndBuffer(fh, buf, ALLOC_SIZE, 0); + auto [file, buffer] = ds->getFileAndBuffer(fh, buf); // TODO: Maintain a data collection to ensure there are no races // on the data @@ -318,8 +318,7 @@ thread_function(int id) idx = stream_dist(gen); auto hip_stream = hip_streams[idx]; - auto [file, buffer, stream] = - ds->getFileBufferAndStream(fh, buf, ALLOC_SIZE, 0, hip_stream); + auto [file, buffer, stream] = ds->getFileBufferAndStream(fh, buf, hip_stream); } break; default: