From 40436e0256acb241b80efb67eaee967a36a2a204 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:05:02 -0600 Subject: [PATCH 1/6] hipFile: Allow setting default implementation for Context. Adds ContextDefaultImpl, which has a type that defaults to the template type. Can be specialized to use different default type for a given template type. --- src/amd_detail/context.h | 11 +++++++++-- 1 file changed, 9 insertions(+), 2 deletions(-) diff --git a/src/amd_detail/context.h b/src/amd_detail/context.h index b568c9f6..19e07ccd 100644 --- a/src/amd_detail/context.h +++ b/src/amd_detail/context.h @@ -6,6 +6,7 @@ #pragma once #include "hipfile-warnings.h" + #include #ifdef AIS_TESTING #include @@ -15,7 +16,13 @@ namespace hipFile { template struct ContextOverride; +template struct ContextDefaultImpl { + using type = T; +}; + template struct Context { + using DefaultImpl = typename ContextDefaultImpl::type; + static_assert(std::is_base_of_v, "ContextDefaultImpl::type must derive from T"); Context() = delete; Context(const Context &) = delete; Context(Context &&) = delete; @@ -32,7 +39,7 @@ template struct Context { { std::lock_guard lock{m}; HIPFILE_WARN_NO_EXIT_DTOR_OFF - static T standard{}; + static DefaultImpl standard{}; HIPFILE_WARN_NO_EXIT_DTOR_ON if (replacement) return replacement; @@ -67,7 +74,7 @@ template struct Context { static T *get() { HIPFILE_WARN_NO_EXIT_DTOR_OFF - static T context{}; + static DefaultImpl context{}; HIPFILE_WARN_NO_EXIT_DTOR_ON return &context; } From 971602a1527487f36d63cdc4307a63e708c7d5a3 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:11:24 -0600 Subject: [PATCH 2/6] hipFile: Create interface for StatsServer. Add an interface for StatsServer, so that a mock doesn't need to call StatsServer constructor. Adds specialization to ContextDefaultImpl so that StatsServer is used as default implementation for IStatsServer. --- src/amd_detail/context.cpp | 2 +- src/amd_detail/context.h | 5 +++++ src/amd_detail/stats.cpp | 32 ++++++++++++++++---------------- src/amd_detail/stats.h | 15 ++++++++++----- test/amd_detail/mstats.h | 4 ++-- 5 files changed, 34 insertions(+), 24 deletions(-) diff --git a/src/amd_detail/context.cpp b/src/amd_detail/context.cpp index 5adc1d87..03dfa838 100644 --- a/src/amd_detail/context.cpp +++ b/src/amd_detail/context.cpp @@ -16,7 +16,7 @@ HipFileInit::HipFileInit() { Context::get(); Context::get(); - Context::get(); + Context::get(); Context::get(); } diff --git a/src/amd_detail/context.h b/src/amd_detail/context.h index 19e07ccd..6b824917 100644 --- a/src/amd_detail/context.h +++ b/src/amd_detail/context.h @@ -6,6 +6,7 @@ #pragma once #include "hipfile-warnings.h" +#include "stats.h" #include #ifdef AIS_TESTING @@ -20,6 +21,10 @@ template struct ContextDefaultImpl { using type = T; }; +template <> struct ContextDefaultImpl { + using type = StatsServer; +}; + template struct Context { using DefaultImpl = typename ContextDefaultImpl::type; static_assert(std::is_base_of_v, "ContextDefaultImpl::type must derive from T"); diff --git a/src/amd_detail/stats.cpp b/src/amd_detail/stats.cpp index f7f722a3..0c19c5c7 100644 --- a/src/amd_detail/stats.cpp +++ b/src/amd_detail/stats.cpp @@ -91,15 +91,25 @@ populateSocketAddr(sockaddr_un &addr, pid_t pid) noexcept } namespace hipFile { +static void +statsDeleter(Stats *s) +{ + if (s == nullptr) { + return; + } + s->~Stats(); + Context::get()->munmap(s, sizeof(Stats)); +} + StatsServer::StatsServer() : m_fd{FileDescriptor::make_managed(Context::get()->memfd_create("AISSTATS", MFD_ALLOW_SEALING))}, - m_efd{FileDescriptor::make_managed(Context::get()->eventfd(0, 0))}, m_stats{nullptr, &statsDeleter} + m_efd{FileDescriptor::make_managed(Context::get()->eventfd(0, 0))}, m_stats{nullptr, statsDeleter} { int fd{m_fd.get()}; Context::get()->ftruncate(fd, sizeof(Stats)); void *shm = Context::get()->mmap(nullptr, sizeof(Stats), PROT_READ | PROT_WRITE, MAP_SHARED, fd, 0); Context::get()->fcntl(fd, F_ADD_SEALS, F_SEAL_SHRINK | F_SEAL_FUTURE_WRITE); - m_stats = UniqueStats{new (shm) Stats{}, &statsDeleter}; + m_stats = std::unique_ptr{new (shm) Stats{}, statsDeleter}; m_stats->level = std::min(static_cast(Context::get()->statsLevel()), StatsLevel::Max); m_thread = std::thread(&StatsServer::threadFn, this); @@ -114,16 +124,6 @@ StatsServer::~StatsServer() } } -void -StatsServer::statsDeleter(Stats *s) -{ - if (s == nullptr) { - return; - } - s->~Stats(); - Context::get()->munmap(s, sizeof(Stats)); -} - void StatsServer::threadFn() { @@ -243,7 +243,7 @@ StatsClient::generateReportV1(std::ostream &stream, const Stats *stats) void statsAddFastPathRead(uint64_t bytes) { - Stats *stats{Context::get()->getStats()}; + Stats *stats{Context::get()->getStats()}; if (stats && stats->level >= StatsLevel::Basic) { stats->getCounter(StatsCounters::TotalFastPathReadBytes) += bytes; } @@ -252,7 +252,7 @@ statsAddFastPathRead(uint64_t bytes) void statsAddFastPathWrite(uint64_t bytes) { - Stats *stats{Context::get()->getStats()}; + Stats *stats{Context::get()->getStats()}; if (stats && stats->level >= StatsLevel::Basic) { stats->getCounter(StatsCounters::TotalFastPathWriteBytes) += bytes; } @@ -261,7 +261,7 @@ statsAddFastPathWrite(uint64_t bytes) void statsAddFallbackPathRead(uint64_t bytes) { - Stats *stats{Context::get()->getStats()}; + Stats *stats{Context::get()->getStats()}; if (stats && stats->level >= StatsLevel::Basic) { stats->getCounter(StatsCounters::TotalFallbackPathReadBytes) += bytes; } @@ -270,7 +270,7 @@ statsAddFallbackPathRead(uint64_t bytes) void statsAddFallbackPathWrite(uint64_t bytes) { - Stats *stats{Context::get()->getStats()}; + Stats *stats{Context::get()->getStats()}; if (stats && stats->level >= StatsLevel::Basic) { stats->getCounter(StatsCounters::TotalFallbackPathWriteBytes) += bytes; } diff --git a/src/amd_detail/stats.h b/src/amd_detail/stats.h index 3351d2fc..217c359f 100644 --- a/src/amd_detail/stats.h +++ b/src/amd_detail/stats.h @@ -55,17 +55,22 @@ void statsAddFastPathWrite(uint64_t bytes); void statsAddFallbackPathRead(uint64_t bytes); void statsAddFallbackPathWrite(uint64_t bytes); -class StatsServer { +class IStatsServer { +public: + virtual ~IStatsServer() = default; + virtual Stats *getStats() = 0; +}; + +class StatsServer : public IStatsServer { public: StatsServer(); - virtual ~StatsServer(); - virtual Stats *getStats() + virtual ~StatsServer() override; + virtual Stats *getStats() override { return m_stats.get(); } - static void statsDeleter(Stats *s); - using UniqueStats = std::unique_ptr; + using UniqueStats = std::unique_ptr; private: void threadFn(); diff --git a/test/amd_detail/mstats.h b/test/amd_detail/mstats.h index 503f5e16..661d97ea 100644 --- a/test/amd_detail/mstats.h +++ b/test/amd_detail/mstats.h @@ -11,8 +11,8 @@ #include namespace hipFile { -class MStatsServer : public StatsServer { - ContextOverride co; +class MStatsServer : public IStatsServer { + ContextOverride co; public: MStatsServer() : co{this} From 8dd8431fb5ab99b6fb145ae0f705833d52096058 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:27:00 -0600 Subject: [PATCH 3/6] hipFile: Defer static init to first API call. Moves current static init to a hipFileInit function. It will be called from any public API function. --- src/amd_detail/context.cpp | 10 ++-------- src/amd_detail/context.h | 9 +++------ src/amd_detail/hipfile.cpp | 29 +++++++++++++++++++++++++++++ test/amd_detail/fallback.cpp | 2 ++ 4 files changed, 36 insertions(+), 14 deletions(-) diff --git a/src/amd_detail/context.cpp b/src/amd_detail/context.cpp index 03dfa838..5bca2028 100644 --- a/src/amd_detail/context.cpp +++ b/src/amd_detail/context.cpp @@ -5,14 +5,14 @@ #include "context.h" #include "hip.h" -#include "hipfile-warnings.h" #include "state.h" #include "stats.h" #include "sys.h" namespace hipFile { -HipFileInit::HipFileInit() +void +hipFileInit() { Context::get(); Context::get(); @@ -20,10 +20,4 @@ HipFileInit::HipFileInit() Context::get(); } -HIPFILE_WARN_NO_GLOBAL_CTOR_OFF -HIPFILE_WARN_NO_EXIT_DTOR_OFF -static HipFileInit *hipfile_init = Context::get(); -HIPFILE_WARN_NO_EXIT_DTOR_ON -HIPFILE_WARN_NO_GLOBAL_CTOR_ON - } diff --git a/src/amd_detail/context.h b/src/amd_detail/context.h index 6b824917..62a46ce4 100644 --- a/src/amd_detail/context.h +++ b/src/amd_detail/context.h @@ -43,11 +43,11 @@ template struct Context { static T *get() { std::lock_guard lock{m}; + if (replacement) + return replacement; HIPFILE_WARN_NO_EXIT_DTOR_OFF static DefaultImpl standard{}; HIPFILE_WARN_NO_EXIT_DTOR_ON - if (replacement) - return replacement; return &standard; } @@ -104,9 +104,6 @@ template struct ContextOverride { }; #endif -class HipFileInit { - HipFileInit(); - friend struct Context; -}; +void hipFileInit(); } diff --git a/src/amd_detail/hipfile.cpp b/src/amd_detail/hipfile.cpp index e54147ff..ff11e6df 100644 --- a/src/amd_detail/hipfile.cpp +++ b/src/amd_detail/hipfile.cpp @@ -69,6 +69,7 @@ ensureDriverInit() hipFileError_t hipFileHandleRegister(hipFileHandle_t *fh, hipFileDescr_t *descr) try { + hipFileInit(); if (fh == nullptr || descr == nullptr) { return {hipFileInvalidValue, hipSuccess}; } @@ -95,6 +96,7 @@ catch (...) { void hipFileHandleDeregister(hipFileHandle_t fh) try { + hipFileInit(); if (fh == nullptr) { return; } @@ -109,6 +111,7 @@ catch (...) { hipFileError_t hipFileBufRegister(const void *buffer_base, size_t length, int flags) try { + hipFileInit(); Context::get()->registerBuffer(buffer_base, length, flags); return {hipFileSuccess, hipSuccess}; } @@ -137,6 +140,7 @@ catch (...) { hipFileError_t hipFileBufDeregister(const void *buffer_base) try { + hipFileInit(); Context::get()->deregisterBuffer(buffer_base); return {hipFileSuccess, hipSuccess}; } @@ -221,6 +225,7 @@ catch (...) { ssize_t hipFileRead(hipFileHandle_t fh, void *buffer_base, size_t size, hoff_t file_offset, hoff_t buffer_offset) { + hipFileInit(); auto result = hipFileIo(IoType::Read, fh, buffer_base, size, file_offset, buffer_offset, getCachedBackends()); @@ -236,6 +241,7 @@ ssize_t hipFileWrite(hipFileHandle_t fh, const void *buffer_base, size_t size, hoff_t file_offset, hoff_t buffer_offset) { + hipFileInit(); auto result = hipFileIo(IoType::Write, fh, buffer_base, size, file_offset, buffer_offset, getCachedBackends()); @@ -250,6 +256,7 @@ hipFileWrite(hipFileHandle_t fh, const void *buffer_base, size_t size, hoff_t fi hipFileError_t hipFileDriverOpen() try { + hipFileInit(); Context::get()->incrRefCount(); return {hipFileSuccess, hipSuccess}; @@ -261,6 +268,7 @@ catch (...) { hipFileError_t hipFileDriverClose() try { + hipFileInit(); if (Context::get()->getRefCount() > 0) { Context::get()->decrRefCount(); return {hipFileSuccess, hipSuccess}; @@ -276,6 +284,7 @@ catch (...) { int64_t hipFileUseCount() try { + hipFileInit(); return Context::get()->getRefCount(); } catch (...) { @@ -285,6 +294,7 @@ catch (...) { hipFileError_t hipFileDriverGetProperties(hipFileDriverProps_t *props) try { + hipFileInit(); (void)props; throw std::runtime_error("Not Implemented"); @@ -296,6 +306,7 @@ catch (...) { hipFileError_t hipFileDriverSetPollMode(bool poll, size_t poll_threshold_size) try { + hipFileInit(); (void)poll; (void)poll_threshold_size; @@ -308,6 +319,7 @@ catch (...) { hipFileError_t hipFileDriverSetMaxDirectIOSize(size_t max_direct_io_size) try { + hipFileInit(); (void)max_direct_io_size; throw std::runtime_error("Not Implemented"); @@ -319,6 +331,7 @@ catch (...) { hipFileError_t hipFileDriverSetMaxCacheSize(size_t max_cache_size) try { + hipFileInit(); (void)max_cache_size; throw std::runtime_error("Not Implemented"); @@ -330,6 +343,7 @@ catch (...) { hipFileError_t hipFileDriverSetMaxPinnedMemSize(size_t max_pinned_size) try { + hipFileInit(); (void)max_pinned_size; throw std::runtime_error("Not Implemented"); @@ -341,6 +355,7 @@ catch (...) { hipFileError_t hipFileBatchIOSetUp(hipFileBatchHandle_t *batch_idp, unsigned max_nr) try { + hipFileInit(); if (batch_idp == nullptr) { return {hipFileInvalidValue, hipSuccess}; } @@ -359,6 +374,7 @@ catch (...) { hipFileError_t hipFileBatchIOSubmit(hipFileBatchHandle_t batch_idp, unsigned nr, hipFileIOParams_t *iocbp, unsigned flags) try { + hipFileInit(); (void)flags; // Unused at this time. std::shared_ptr batch_context = Context::get()->getBatchContext(batch_idp); @@ -377,6 +393,7 @@ hipFileError_t hipFileBatchIOGetStatus(hipFileBatchHandle_t batch_idp, unsigned min_nr, unsigned *nr, hipFileIOEvents_t *iocbp, struct timespec *timeout) try { + hipFileInit(); (void)batch_idp; (void)min_nr; (void)nr; @@ -392,6 +409,7 @@ catch (...) { hipFileError_t hipFileBatchIOCancel(hipFileBatchHandle_t batch_idp) try { + hipFileInit(); (void)batch_idp; throw std::runtime_error("Not Implemented"); @@ -403,6 +421,7 @@ catch (...) { void hipFileBatchIODestroy(hipFileBatchHandle_t batch_idp) try { + hipFileInit(); (void)batch_idp; throw std::runtime_error("Not Implemented"); @@ -449,6 +468,7 @@ hipFileError_t hipFileReadAsync(hipFileHandle_t fh, void *buffer_base, size_t *size_p, hoff_t *file_offset_p, hoff_t *buffer_offset_p, ssize_t *bytes_read_p, hipStream_t stream) { + hipFileInit(); return hipFileIOAsync(IoType::Read, fh, buffer_base, size_p, file_offset_p, buffer_offset_p, bytes_read_p, stream); } @@ -457,6 +477,7 @@ hipFileError_t hipFileWriteAsync(hipFileHandle_t fh, void *buffer_base, size_t *size_p, hoff_t *file_offset_p, hoff_t *buffer_offset_p, ssize_t *bytes_written_p, hipStream_t stream) { + hipFileInit(); return hipFileIOAsync(IoType::Write, fh, buffer_base, size_p, file_offset_p, buffer_offset_p, bytes_written_p, stream); } @@ -464,6 +485,7 @@ hipFileWriteAsync(hipFileHandle_t fh, void *buffer_base, size_t *size_p, hoff_t hipFileError_t hipFileStreamRegister(hipStream_t stream, unsigned flags) try { + hipFileInit(); Context::get()->registerStream(stream, flags); return {hipFileSuccess, hipSuccess}; } @@ -477,6 +499,7 @@ catch (...) { hipFileError_t hipFileStreamDeregister(hipStream_t stream) try { + hipFileInit(); Context::get()->deregisterStream(stream); return {hipFileSuccess, hipSuccess}; } @@ -494,6 +517,7 @@ catch (...) { hipFileError_t hipFileGetParameterSizeT(hipFileSizeTConfigParameter_t param, size_t *value) try { + hipFileInit(); (void)param; (void)value; @@ -506,6 +530,7 @@ catch (...) { hipFileError_t hipFileGetParameterBool(hipFileBoolConfigParameter_t param, bool *value) try { + hipFileInit(); (void)param; (void)value; @@ -518,6 +543,7 @@ catch (...) { hipFileError_t hipFileGetParameterString(hipFileStringConfigParameter_t param, char *desc_str, int len) try { + hipFileInit(); (void)param; (void)desc_str; (void)len; @@ -531,6 +557,7 @@ catch (...) { hipFileError_t hipFileSetParameterSizeT(hipFileSizeTConfigParameter_t param, size_t value) try { + hipFileInit(); (void)param; (void)value; @@ -543,6 +570,7 @@ catch (...) { hipFileError_t hipFileSetParameterBool(hipFileBoolConfigParameter_t param, bool value) try { + hipFileInit(); (void)param; (void)value; @@ -555,6 +583,7 @@ catch (...) { hipFileError_t hipFileSetParameterString(hipFileStringConfigParameter_t param, const char *desc_str) try { + hipFileInit(); (void)param; (void)desc_str; diff --git a/test/amd_detail/fallback.cpp b/test/amd_detail/fallback.cpp index f257727a..7791f96d 100644 --- a/test/amd_detail/fallback.cpp +++ b/test/amd_detail/fallback.cpp @@ -18,6 +18,7 @@ #include "mfile.h" #include "mhip.h" #include "mmountinfo.h" +#include "mstats.h" #include "msys.h" #include "state.h" @@ -183,6 +184,7 @@ struct FallbackParam : ::testing::TestWithParam { StrictMock mhip{}; StrictMock msys{}; + StrictMock mserver{}; StrictMock mlibmounthelper{}; StrictMock mcfg{}; From fcb760926672b138202bdb048bd24ee98094658d Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:30:28 -0600 Subject: [PATCH 4/6] hipFile: Add system test to test hipFileRead in new thread. --- test/system/amd/io.cpp | 59 ++++++++++++++++++++++++++++++++++++++++++ 1 file changed, 59 insertions(+) diff --git a/test/system/amd/io.cpp b/test/system/amd/io.cpp index 45200110..e51b4e3a 100644 --- a/test/system/amd/io.cpp +++ b/test/system/amd/io.cpp @@ -16,6 +16,7 @@ #include #include #include +#include #include extern SystemTestOptions test_env; @@ -114,4 +115,62 @@ INSTANTIATE_TEST_SUITE_P(, HipFileIo, testing::ValuesIn(io_test_params), return param_info.param.name; }); +struct HipFileIoHipInit : public testing::Test { + + Tmpfile tmpfile; + size_t tmpfile_size; + hipFileHandle_t tmpfile_handle; + void *registered_device_buffer; + size_t registered_device_buffer_size; + + HipFileIoHipInit() + : tmpfile{test_env.ais_capable_dir}, tmpfile_size{1024 * 1024}, tmpfile_handle{nullptr}, + registered_device_buffer{nullptr}, registered_device_buffer_size{1024 * 1024} + { + } + + // Must be called before hipfile is initialized. Relies on each test being + // run in a separate process + void enable_fastpath_only() + { + if (unsetenv("HIPFILE_FORCE_COMPAT_MODE")) { + FAIL() << "Could not clear HIPFILE_FORCE_COMPAT_MODE"; + } + if (setenv("HIPFILE_ALLOW_COMPAT_MODE", "false", 1)) { + FAIL() << "Could not set HIPFILE_ALLOW_COMPAT_MODE=false"; + } + } + + void SetUp() override + { + enable_fastpath_only(); + + 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(®istered_device_buffer, registered_device_buffer_size)); + ASSERT_EQ(HIPFILE_SUCCESS, + hipFileBufRegister(registered_device_buffer, registered_device_buffer_size, 0)); + } + + void TearDown() override + { + ASSERT_EQ(HIPFILE_SUCCESS, hipFileBufDeregister(registered_device_buffer)); + ASSERT_EQ(hipSuccess, hipFree(registered_device_buffer)); + hipFileHandleDeregister(tmpfile_handle); + } +}; + +TEST_F(HipFileIoHipInit, spawnedThreadOperationRunsWithoutSegfault) +{ + size_t io_size{registered_device_buffer_size}; + std::thread([&]() { + ASSERT_EQ(io_size, hipFileRead(tmpfile_handle, registered_device_buffer, io_size, 0, 0)); + }).join(); +} + HIPFILE_WARN_NO_GLOBAL_CTOR_ON From 4a7281397fbf43cbb63f316e6be41f09cfc4f452 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 11:25:50 -0600 Subject: [PATCH 5/6] NEED TO SQUASH: Updates to tests after rebase on develop. --- test/amd_detail/fallback.cpp | 3 +++ test/amd_detail/fastpath.cpp | 6 ++++++ 2 files changed, 9 insertions(+) diff --git a/test/amd_detail/fallback.cpp b/test/amd_detail/fallback.cpp index 7791f96d..a0f99880 100644 --- a/test/amd_detail/fallback.cpp +++ b/test/amd_detail/fallback.cpp @@ -268,6 +268,7 @@ TEST_P(FallbackParam, FallbackIoTruncatesSizeToMAX_RW_COUNT) EXPECT_CALL(msys, mmap).WillOnce(testing::Return(reinterpret_cast(0xFEFEFEFE))); switch (io_type) { case IoType::Read: + EXPECT_CALL(mserver, getStats); EXPECT_CALL(msys, pread) .WillRepeatedly(testing::Invoke([](int, void *, size_t count, hoff_t) -> ssize_t { return static_cast(count); @@ -275,6 +276,7 @@ TEST_P(FallbackParam, FallbackIoTruncatesSizeToMAX_RW_COUNT) EXPECT_CALL(mhip, hipMemcpy).WillRepeatedly(testing::Return()); break; case IoType::Write: + EXPECT_CALL(mserver, getStats); EXPECT_CALL(mhip, hipMemcpy).WillRepeatedly(testing::Return()); EXPECT_CALL(mhip, hipStreamSynchronize).WillRepeatedly(testing::Return()); EXPECT_CALL(msys, pwrite) @@ -306,6 +308,7 @@ TEST_P(FallbackParam, FallbackIoAllocatesChunkSizedHostBounceBuffer) EXPECT_CALL(mcfg, fallback()).WillOnce(Return(true)); EXPECT_CALL(msys, mmap(testing::_, chunk_size, testing::_, testing::_, testing::_, testing::_)) .WillOnce(testing::Return(ptr)); + EXPECT_CALL(mserver, getStats); switch (io_type) { case IoType::Read: EXPECT_CALL(msys, pread).WillOnce(testing::Return(0)); diff --git a/test/amd_detail/fastpath.cpp b/test/amd_detail/fastpath.cpp index 06098b70..52c1a913 100644 --- a/test/amd_detail/fastpath.cpp +++ b/test/amd_detail/fastpath.cpp @@ -442,6 +442,7 @@ TEST_P(FastpathIoParam, IoConfiguresHandle) handle.fd = DEFAULT_UNBUFFERED_FD.value(); expect_io(); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(Eq(handle), _, _, _)); @@ -463,6 +464,7 @@ TEST_P(FastpathIoParam, IoCalculatesCorrectDevicePointer) void *expected_device_ptr{reinterpret_cast(0x21000)}; expect_io(DEFAULT_UNBUFFERED_FD, buffer_addr, static_cast(buffer_offset) + DEFAULT_IO_SIZE); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, expected_device_ptr, _, _)); @@ -480,6 +482,7 @@ TEST_P(FastpathIoParam, IoCalculatesCorrectDevicePointer) TEST_P(FastpathIoParam, IoPassesThroughSizeAndFileOffset) { expect_io(); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, _, Eq(DEFAULT_IO_SIZE), Eq(DEFAULT_FILE_OFFSET))); @@ -497,6 +500,7 @@ TEST_P(FastpathIoParam, IoPassesThroughSizeAndFileOffset) TEST_P(FastpathIoParam, IoReturnsBytesTransferred) { expect_io(); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, _, _, _)).WillOnce(Return(DEFAULT_IO_SIZE)); @@ -520,6 +524,7 @@ TEST_P(FastpathIoParam, IoReturnsBytesTransferredShort) ASSERT_LT(nbytes, DEFAULT_IO_SIZE); expect_io(); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, _, _, _)).WillOnce(Return(nbytes)); @@ -583,6 +588,7 @@ TEST_P(FastpathIoParam, IoSizeIsTruncatedToMaxRWCount) const size_t io_size{SIZE_MAX}; expect_io(DEFAULT_UNBUFFERED_FD, DEFAULT_BUFFER_ADDR, buffer_size); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, _, MAX_RW_COUNT, _)).WillOnce(Return(MAX_RW_COUNT)); From 30811a5137d4be7eb3f901713a11a682ab077ea2 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 11:49:34 -0600 Subject: [PATCH 6/6] hipFile: Remove HIPFILE_STATICs. --- src/amd_detail/configuration.cpp | 10 +++++----- src/amd_detail/hip.cpp | 4 ++-- 2 files changed, 7 insertions(+), 7 deletions(-) diff --git a/src/amd_detail/configuration.cpp b/src/amd_detail/configuration.cpp index 97b88e19..89374f18 100644 --- a/src/amd_detail/configuration.cpp +++ b/src/amd_detail/configuration.cpp @@ -15,9 +15,9 @@ using namespace hipFile; bool Configuration::fastpath() const noexcept { - HIPFILE_STATIC bool fastpath_env{!Environment::force_compat_mode().value_or(false)}; - HIPFILE_STATIC bool readExists{!!getHipAmdFileReadPtr()}; - HIPFILE_STATIC bool writeExists{!!getHipAmdFileWritePtr()}; + bool fastpath_env{!Environment::force_compat_mode().value_or(false)}; + bool readExists{!!getHipAmdFileReadPtr()}; + bool writeExists{!!getHipAmdFileWritePtr()}; return readExists && writeExists && m_fastpath_override.value_or(fastpath_env); } @@ -30,7 +30,7 @@ Configuration::fastpath(bool enabled) noexcept bool Configuration::fallback() const noexcept { - HIPFILE_STATIC bool fallback_env{Environment::allow_compat_mode().value_or(true)}; + bool fallback_env{Environment::allow_compat_mode().value_or(true)}; return m_fallback_override.value_or(fallback_env); } @@ -43,6 +43,6 @@ Configuration::fallback(bool enabled) noexcept unsigned int Configuration::statsLevel() const noexcept { - HIPFILE_STATIC unsigned int stats_level_env{Environment::stats_level().value_or(0)}; + unsigned int stats_level_env{Environment::stats_level().value_or(0)}; return stats_level_env; } diff --git a/src/amd_detail/hip.cpp b/src/amd_detail/hip.cpp index 66a9f149..1b018016 100644 --- a/src/amd_detail/hip.cpp +++ b/src/amd_detail/hip.cpp @@ -25,7 +25,7 @@ catch (...) { hipAmdFileRead_t getHipAmdFileReadPtr() { - HIPFILE_STATIC hipAmdFileRead_t hipAmdFileReadPtr{ + hipAmdFileRead_t hipAmdFileReadPtr{ reinterpret_cast(hipGetProcAddressHelper("hipAmdFileRead"))}; return hipAmdFileReadPtr; } @@ -33,7 +33,7 @@ getHipAmdFileReadPtr() hipAmdFileWrite_t getHipAmdFileWritePtr() { - HIPFILE_STATIC hipAmdFileWrite_t hipAmdFileWritePtr{ + hipAmdFileWrite_t hipAmdFileWritePtr{ reinterpret_cast(hipGetProcAddressHelper("hipAmdFileWrite"))}; return hipAmdFileWritePtr; }