From 74c12493032cb49b648fda6247b7e4e412e86316 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:05:02 -0600 Subject: [PATCH 1/8] context: 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 f71e5d915e0d21c75ee3bef8f888d7e4e4b88aa7 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:11:24 -0600 Subject: [PATCH 2/8] stats: 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 | 8 ++++---- src/amd_detail/stats.h | 12 +++++++++--- test/amd_detail/mstats.h | 4 ++-- 5 files changed, 21 insertions(+), 10 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 466fb840..a557ebc5 100644 --- a/src/amd_detail/stats.cpp +++ b/src/amd_detail/stats.cpp @@ -262,7 +262,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; } @@ -271,7 +271,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; } @@ -280,7 +280,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; } @@ -289,7 +289,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 c4a3bc94..65802d82 100644 --- a/src/amd_detail/stats.h +++ b/src/amd_detail/stats.h @@ -55,11 +55,17 @@ 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(); } 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 deaa84913754f6e45bafd109c16dea444600bdfa Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Mon, 6 Apr 2026 10:58:01 -0600 Subject: [PATCH 3/8] context: Return replacement before initializing standard implementation. If multiple interfaces are being mocked, the standard implementation may end up calling mocked functions when constructing. Return the replacement first, so we only construct the standard implementation when it is actually needed. --- src/amd_detail/context.h | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/src/amd_detail/context.h b/src/amd_detail/context.h index 6b824917..bff21ba4 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; } From a56c011c1c1f02090febd47e242d84ddb7cfb266 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:27:00 -0600 Subject: [PATCH 4/8] 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 | 5 +--- src/amd_detail/hipfile.cpp | 51 +++++++++++++++++++++++++++++++++--- test/amd_detail/fallback.cpp | 5 ++++ test/amd_detail/fastpath.cpp | 6 +++++ 5 files changed, 61 insertions(+), 16 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 bff21ba4..62a46ce4 100644 --- a/src/amd_detail/context.h +++ b/src/amd_detail/context.h @@ -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..2d748165 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}; } @@ -220,7 +224,8 @@ catch (...) { ssize_t hipFileRead(hipFileHandle_t fh, void *buffer_base, size_t size, hoff_t file_offset, hoff_t buffer_offset) -{ +try { + hipFileInit(); auto result = hipFileIo(IoType::Read, fh, buffer_base, size, file_offset, buffer_offset, getCachedBackends()); @@ -231,11 +236,16 @@ hipFileRead(hipFileHandle_t fh, void *buffer_base, size_t size, hoff_t file_offs } return result; } +catch (...) { + hipFileError_t err = handle_exception(); + return -err.err; +} ssize_t hipFileWrite(hipFileHandle_t fh, const void *buffer_base, size_t size, hoff_t file_offset, hoff_t buffer_offset) -{ +try { + hipFileInit(); auto result = hipFileIo(IoType::Write, fh, buffer_base, size, file_offset, buffer_offset, getCachedBackends()); @@ -246,10 +256,15 @@ hipFileWrite(hipFileHandle_t fh, const void *buffer_base, size_t size, hoff_t fi } return result; } +catch (...) { + hipFileError_t err = handle_exception(); + return -err.err; +} hipFileError_t hipFileDriverOpen() try { + hipFileInit(); Context::get()->incrRefCount(); return {hipFileSuccess, hipSuccess}; @@ -261,6 +276,7 @@ catch (...) { hipFileError_t hipFileDriverClose() try { + hipFileInit(); if (Context::get()->getRefCount() > 0) { Context::get()->decrRefCount(); return {hipFileSuccess, hipSuccess}; @@ -276,6 +292,7 @@ catch (...) { int64_t hipFileUseCount() try { + hipFileInit(); return Context::get()->getRefCount(); } catch (...) { @@ -285,6 +302,7 @@ catch (...) { hipFileError_t hipFileDriverGetProperties(hipFileDriverProps_t *props) try { + hipFileInit(); (void)props; throw std::runtime_error("Not Implemented"); @@ -296,6 +314,7 @@ catch (...) { hipFileError_t hipFileDriverSetPollMode(bool poll, size_t poll_threshold_size) try { + hipFileInit(); (void)poll; (void)poll_threshold_size; @@ -308,6 +327,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 +339,7 @@ catch (...) { hipFileError_t hipFileDriverSetMaxCacheSize(size_t max_cache_size) try { + hipFileInit(); (void)max_cache_size; throw std::runtime_error("Not Implemented"); @@ -330,6 +351,7 @@ catch (...) { hipFileError_t hipFileDriverSetMaxPinnedMemSize(size_t max_pinned_size) try { + hipFileInit(); (void)max_pinned_size; throw std::runtime_error("Not Implemented"); @@ -341,6 +363,7 @@ catch (...) { hipFileError_t hipFileBatchIOSetUp(hipFileBatchHandle_t *batch_idp, unsigned max_nr) try { + hipFileInit(); if (batch_idp == nullptr) { return {hipFileInvalidValue, hipSuccess}; } @@ -359,6 +382,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 +401,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 +417,7 @@ catch (...) { hipFileError_t hipFileBatchIOCancel(hipFileBatchHandle_t batch_idp) try { + hipFileInit(); (void)batch_idp; throw std::runtime_error("Not Implemented"); @@ -403,6 +429,7 @@ catch (...) { void hipFileBatchIODestroy(hipFileBatchHandle_t batch_idp) try { + hipFileInit(); (void)batch_idp; throw std::runtime_error("Not Implemented"); @@ -448,22 +475,31 @@ catch (...) { 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) -{ +try { + hipFileInit(); return hipFileIOAsync(IoType::Read, fh, buffer_base, size_p, file_offset_p, buffer_offset_p, bytes_read_p, stream); } +catch (...) { + return handle_exception(); +} 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) -{ +try { + hipFileInit(); return hipFileIOAsync(IoType::Write, fh, buffer_base, size_p, file_offset_p, buffer_offset_p, bytes_written_p, stream); } +catch (...) { + return handle_exception(); +} hipFileError_t hipFileStreamRegister(hipStream_t stream, unsigned flags) try { + hipFileInit(); Context::get()->registerStream(stream, flags); return {hipFileSuccess, hipSuccess}; } @@ -477,6 +513,7 @@ catch (...) { hipFileError_t hipFileStreamDeregister(hipStream_t stream) try { + hipFileInit(); Context::get()->deregisterStream(stream); return {hipFileSuccess, hipSuccess}; } @@ -494,6 +531,7 @@ catch (...) { hipFileError_t hipFileGetParameterSizeT(hipFileSizeTConfigParameter_t param, size_t *value) try { + hipFileInit(); (void)param; (void)value; @@ -506,6 +544,7 @@ catch (...) { hipFileError_t hipFileGetParameterBool(hipFileBoolConfigParameter_t param, bool *value) try { + hipFileInit(); (void)param; (void)value; @@ -518,6 +557,7 @@ catch (...) { hipFileError_t hipFileGetParameterString(hipFileStringConfigParameter_t param, char *desc_str, int len) try { + hipFileInit(); (void)param; (void)desc_str; (void)len; @@ -531,6 +571,7 @@ catch (...) { hipFileError_t hipFileSetParameterSizeT(hipFileSizeTConfigParameter_t param, size_t value) try { + hipFileInit(); (void)param; (void)value; @@ -543,6 +584,7 @@ catch (...) { hipFileError_t hipFileSetParameterBool(hipFileBoolConfigParameter_t param, bool value) try { + hipFileInit(); (void)param; (void)value; @@ -555,6 +597,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 0481dbd0..c48f42b1 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{}; @@ -266,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); @@ -273,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) @@ -305,6 +309,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 013b329d..72320777 100644 --- a/test/amd_detail/fastpath.cpp +++ b/test/amd_detail/fastpath.cpp @@ -537,6 +537,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), _, _, _)); @@ -558,6 +559,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, _, _)); @@ -575,6 +577,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))); @@ -592,6 +595,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)); @@ -615,6 +619,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)); @@ -678,6 +683,7 @@ TEST_P(FastpathIoParam, IoSizeIsTruncatedToMaxRWCount) const size_t io_size{SIZE_MAX}; expect_io(DEFAULT_UNBUFFERED_FD, reinterpret_cast(DEFAULT_BUFFER_ADDR), buffer_size); + EXPECT_CALL(mcfg, statsLevel); switch (GetParam()) { case IoType::Read: EXPECT_CALL(mhip, hipAmdFileRead(_, _, getMaxRwCount(), _)).WillOnce(Return(getMaxRwCount())); From 2cd03e357dc1cc3796d015996c77ec9301166ca8 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Tue, 24 Mar 2026 10:30:28 -0600 Subject: [PATCH 5/8] io: Add system tests to test hipFileRead/Write in new thread. --- test/system/amd/io.cpp | 58 ++++++++++++++++++++++++++++++++++++++++++ 1 file changed, 58 insertions(+) diff --git a/test/system/amd/io.cpp b/test/system/amd/io.cpp index 45200110..4d8dd0e3 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,61 @@ 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} + { + } + + void SetUp() override + { + Context::get()->fastpath(true); + Context::get()->fallback(false); + + 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, spawnedThreadReadRunsWithoutSegfault) +{ + size_t io_size{registered_device_buffer_size}; + ssize_t res{}; + std::thread([&]() { res = hipFileRead(tmpfile_handle, registered_device_buffer, io_size, 0, 0); }).join(); + ASSERT_EQ(io_size, res); +} + +TEST_F(HipFileIoHipInit, spawnedThreadWriteRunsWithoutSegfault) +{ + size_t io_size{registered_device_buffer_size}; + ssize_t res{}; + std::thread([&]() { + res = hipFileWrite(tmpfile_handle, registered_device_buffer, io_size, 0, 0); + }).join(); + ASSERT_EQ(io_size, res); +} + HIPFILE_WARN_NO_GLOBAL_CTOR_ON From 7f4fccf89dd31f1036039d32e8d58426200366cf Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Mon, 6 Apr 2026 17:42:20 -0400 Subject: [PATCH 6/8] config,hip: Revert HIPFILE_STATICs back to statics. --- src/amd_detail/configuration.cpp | 13 ++++++------- src/amd_detail/hip.cpp | 5 ++--- src/amd_detail/static.h | 15 --------------- test/amd_detail/configuration.cpp | 6 ------ 4 files changed, 8 insertions(+), 31 deletions(-) delete mode 100644 src/amd_detail/static.h diff --git a/src/amd_detail/configuration.cpp b/src/amd_detail/configuration.cpp index 427e8622..82a1cb70 100644 --- a/src/amd_detail/configuration.cpp +++ b/src/amd_detail/configuration.cpp @@ -6,7 +6,6 @@ #include "configuration.h" #include "environment.h" #include "hip.h" -#include "static.h" #include @@ -15,9 +14,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()}; + static bool fastpath_env{!Environment::force_compat_mode().value_or(false)}; + static bool readExists{!!getHipAmdFileReadPtr()}; + static bool writeExists{!!getHipAmdFileWritePtr()}; return readExists && writeExists && m_fastpath_override.value_or(fastpath_env); } @@ -30,7 +29,7 @@ Configuration::fastpath(bool enabled) noexcept bool Configuration::fallback() const noexcept { - HIPFILE_STATIC bool fallback_env{Environment::allow_compat_mode().value_or(true)}; + static bool fallback_env{Environment::allow_compat_mode().value_or(true)}; return m_fallback_override.value_or(fallback_env); } @@ -43,13 +42,13 @@ Configuration::fallback(bool enabled) noexcept unsigned int Configuration::statsLevel() const noexcept { - HIPFILE_STATIC unsigned int stats_level_env{Environment::stats_level().value_or(0)}; + static unsigned int stats_level_env{Environment::stats_level().value_or(0)}; return stats_level_env; } bool Configuration::unsupportedFileSystems() const noexcept { - HIPFILE_STATIC bool unsupported_file_systems_env{Environment::unsupported_file_systems().value_or(false)}; + static bool unsupported_file_systems_env{Environment::unsupported_file_systems().value_or(false)}; return unsupported_file_systems_env; } diff --git a/src/amd_detail/hip.cpp b/src/amd_detail/hip.cpp index 66a9f149..e009fe6d 100644 --- a/src/amd_detail/hip.cpp +++ b/src/amd_detail/hip.cpp @@ -5,7 +5,6 @@ #include "context.h" #include "hip.h" -#include "static.h" #include #include @@ -25,7 +24,7 @@ catch (...) { hipAmdFileRead_t getHipAmdFileReadPtr() { - HIPFILE_STATIC hipAmdFileRead_t hipAmdFileReadPtr{ + static hipAmdFileRead_t hipAmdFileReadPtr{ reinterpret_cast(hipGetProcAddressHelper("hipAmdFileRead"))}; return hipAmdFileReadPtr; } @@ -33,7 +32,7 @@ getHipAmdFileReadPtr() hipAmdFileWrite_t getHipAmdFileWritePtr() { - HIPFILE_STATIC hipAmdFileWrite_t hipAmdFileWritePtr{ + static hipAmdFileWrite_t hipAmdFileWritePtr{ reinterpret_cast(hipGetProcAddressHelper("hipAmdFileWrite"))}; return hipAmdFileWritePtr; } diff --git a/src/amd_detail/static.h b/src/amd_detail/static.h deleted file mode 100644 index ca0b14d6..00000000 --- a/src/amd_detail/static.h +++ /dev/null @@ -1,15 +0,0 @@ -/* Copyright (c) Advanced Micro Devices, Inc. All rights reserved. - * - * SPDX-License-Identifier: MIT - */ - -#pragma once - -// When testing it is sometimes inconvenient if a variable is statically initialized. -// Declaring a variable as STATIC will result in the variable being static only -// when tests are not being built. -#ifdef AIS_TESTING -#define HIPFILE_STATIC -#else -#define HIPFILE_STATIC static -#endif diff --git a/test/amd_detail/configuration.cpp b/test/amd_detail/configuration.cpp index 3a719940..282b2bea 100644 --- a/test/amd_detail/configuration.cpp +++ b/test/amd_detail/configuration.cpp @@ -81,7 +81,6 @@ TEST_F(HipFileConfiguration, OverrideEnabledFastpathBackend) ASSERT_TRUE(config.fastpath()); config.fastpath(false); - expect_configuration_fastpath("false"); ASSERT_FALSE(config.fastpath()); } @@ -110,7 +109,6 @@ TEST_F(HipFileConfiguration, OverrideDisabledFastpathBackend) ASSERT_FALSE(config.fastpath()); config.fastpath(true); - expect_configuration_fastpath("true"); ASSERT_TRUE(config.fastpath()); } @@ -121,7 +119,6 @@ TEST_F(HipFileConfiguration, CantOverrideDisabledFastpathBackendIfHipAmdFileRead ASSERT_FALSE(config.fastpath()); config.fastpath(true); - expect_configuration_fastpath(nullptr, nullptr); ASSERT_FALSE(config.fastpath()); } @@ -132,7 +129,6 @@ TEST_F(HipFileConfiguration, CantOverrideDisabledFastpathBackendIfHipAmdFileWrit ASSERT_FALSE(config.fastpath()); config.fastpath(true); - expect_configuration_fastpath(nullptr, reinterpret_cast(0x1), nullptr); ASSERT_FALSE(config.fastpath()); } @@ -161,7 +157,6 @@ TEST_F(HipFileConfiguration, OverrideEnabledFallbackBackend) ASSERT_TRUE(config.fallback()); config.fallback(false); - expect_configuration_fallback(nullptr); ASSERT_FALSE(config.fallback()); } @@ -178,7 +173,6 @@ TEST_F(HipFileConfiguration, OverrideDisabledFallbackBackend) ASSERT_FALSE(config.fallback()); config.fallback(true); - expect_configuration_fallback("false"); ASSERT_TRUE(config.fallback()); } From 90e0c95514ea740d6583f55d22796a1467db3cc3 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Wed, 25 Mar 2026 16:27:38 -0600 Subject: [PATCH 7/8] context: Add conditionally compiled requires constraint. Allows use when compiled in C++20 for better type checking, but left out when compiled in C++17 mode. --- CMakeLists.txt | 3 +++ cmake/AISAddExecutable.cmake | 2 +- cmake/AISAddLibraries.cmake | 2 +- shared/hipfile-cpp20.h | 13 +++++++++++++ src/amd_detail/context.h | 18 +++++++++++------- 5 files changed, 29 insertions(+), 9 deletions(-) create mode 100644 shared/hipfile-cpp20.h diff --git a/CMakeLists.txt b/CMakeLists.txt index c51c1fc8..95c5c25a 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -47,6 +47,9 @@ set(HIPFILE_SRC_COMMON_PATH "${CMAKE_CURRENT_SOURCE_DIR}/src/common") set(HIPFILE_TEST_COMMON_PATH "${CMAKE_CURRENT_SOURCE_DIR}/test/common") set(HIPFILE_AMD_TEST_PATH "${CMAKE_CURRENT_SOURCE_DIR}/test/amd_detail") +set(AIS_CXX_STANDARD "17" CACHE STRING "C++ standard to build with") +set_property(CACHE AIS_CXX_STANDARD PROPERTY STRINGS "17" "20") + # Set the list of build types for ccmake/CMake GUI # # This turns the build type box into a multi-select so you don't have to diff --git a/cmake/AISAddExecutable.cmake b/cmake/AISAddExecutable.cmake index 15af7024..bb252ee0 100644 --- a/cmake/AISAddExecutable.cmake +++ b/cmake/AISAddExecutable.cmake @@ -30,7 +30,7 @@ function(ais_add_executable) ais_set_compiler_flags(${arg_NAME}) # Set C++ standard - target_compile_features(${arg_NAME} PRIVATE cxx_std_17) + target_compile_features(${arg_NAME} PRIVATE cxx_std_${AIS_CXX_STANDARD}) set_target_properties(${arg_NAME} PROPERTIES CXX_EXTENSIONS OFF) # Turn sanitizers off for executables diff --git a/cmake/AISAddLibraries.cmake b/cmake/AISAddLibraries.cmake index f40cb995..a7c0791c 100644 --- a/cmake/AISAddLibraries.cmake +++ b/cmake/AISAddLibraries.cmake @@ -28,7 +28,7 @@ function(ais_add_libraries) add_library(${arg_NAME} ${arg_SRCS}) # Set C++ standard - target_compile_features(${arg_NAME} PUBLIC cxx_std_17) + target_compile_features(${arg_NAME} PUBLIC cxx_std_${AIS_CXX_STANDARD}) set_target_properties(${arg_NAME} PROPERTIES CXX_EXTENSIONS OFF) # Set position-independent code diff --git a/shared/hipfile-cpp20.h b/shared/hipfile-cpp20.h new file mode 100644 index 00000000..67449d01 --- /dev/null +++ b/shared/hipfile-cpp20.h @@ -0,0 +1,13 @@ +/* Copyright (c) Advanced Micro Devices, Inc. All rights reserved. + * + * SPDX-License-Identifier: MIT + */ + +#pragma once + +#if __cplusplus >= 202002L +#include +#define HIPFILE_REQUIRES(...) requires(__VA_ARGS__) +#else +#define HIPFILE_REQUIRES(...) +#endif diff --git a/src/amd_detail/context.h b/src/amd_detail/context.h index 62a46ce4..fd9c89df 100644 --- a/src/amd_detail/context.h +++ b/src/amd_detail/context.h @@ -5,6 +5,7 @@ #pragma once +#include "hipfile-cpp20.h" #include "hipfile-warnings.h" #include "stats.h" @@ -13,21 +14,24 @@ #include #endif +#define HIPFILE_CONTEXT_DEFAULT_IMPL(T, Impl) \ + template <> struct ContextDefaultImpl : ContextDefaultImplChecked {} + namespace hipFile { template struct ContextOverride; -template struct ContextDefaultImpl { - using type = T; +template +HIPFILE_REQUIRES(std::derived_from) +struct ContextDefaultImplChecked { + using type = Impl; }; -template <> struct ContextDefaultImpl { - using type = StatsServer; -}; +template struct ContextDefaultImpl : ContextDefaultImplChecked {}; +HIPFILE_CONTEXT_DEFAULT_IMPL(IStatsServer, StatsServer); template struct Context { - using DefaultImpl = typename ContextDefaultImpl::type; - static_assert(std::is_base_of_v, "ContextDefaultImpl::type must derive from T"); + using DefaultImpl = typename ContextDefaultImpl::type; Context() = delete; Context(const Context &) = delete; Context(Context &&) = delete; From 7437235545638bd02de342bbf85fea42c2d75be4 Mon Sep 17 00:00:00 2001 From: Jordan Patterson Date: Mon, 6 Apr 2026 11:30:12 -0600 Subject: [PATCH 8/8] ci: Add a amdclang++ C++20 build. --- .github/workflows/ais-ci.yml | 12 ++++++++++++ .github/workflows/build-ais.yml | 17 ++++++++++++++--- 2 files changed, 26 insertions(+), 3 deletions(-) diff --git a/.github/workflows/ais-ci.yml b/.github/workflows/ais-ci.yml index b323b741..5dcdf27e 100644 --- a/.github/workflows/ais-ci.yml +++ b/.github/workflows/ais-ci.yml @@ -48,6 +48,18 @@ jobs: platform: ${{ inputs.platform }} rocm_version: ${{ inputs.rocm_version }} upload_artifacts: true + build_and_test_cxx20: + if: ${{ !cancelled() && needs.build_AIS_CI_image.outputs.ci_image_build_satisfied == 'true' }} + needs: [AIS_CI_Pre-check, build_AIS_CI_image] + uses: ./.github/workflows/build-ais.yml + with: + ci_image: ${{ needs.build_AIS_CI_image.outputs.ci_image }} + cxx_compiler: amdclang++ + cxx_standard: 20 + platform: ${{ inputs.platform }} + rocm_version: ${{ inputs.rocm_version }} + upload_artifacts: false + generate_coverage: false build_and_test_other_compilers: # Try building on other compilers, but keep as separate job. # Removes issues regarding matrix job overwriting outputs and diff --git a/.github/workflows/build-ais.yml b/.github/workflows/build-ais.yml index c86a2e9a..2b56d1d3 100644 --- a/.github/workflows/build-ais.yml +++ b/.github/workflows/build-ais.yml @@ -5,6 +5,7 @@ env: AIS_INPUT_BUILD_ID: ${{ inputs.build_id }} AIS_INPUT_CI_IMAGE: ${{ inputs.ci_image }} AIS_INPUT_CXX_COMPILER: ${{ inputs.cxx_compiler }} + AIS_INPUT_CXX_STANDARD: ${{ inputs.cxx_standard }} AIS_INPUT_JOB_DESIGNATOR: ${{ inputs.job_designator }} AIS_INPUT_PLATFORM: ${{ inputs.platform }} AIS_INPUT_UPLOAD_ARTIFACTS: ${{ inputs.upload_artifacts }} @@ -18,7 +19,7 @@ env: AIS_PKG_TYPE: ${{ inputs.platform == 'ubuntu' && 'DEB' || 'RPM' }} AIS_INPUT_ROCM_VERSION: ${{ inputs.rocm_version }} # Code coverage report only vetted to work for amdclang++ on Ubuntu - AIS_USE_CODE_COVERAGE: ${{ inputs.cxx_compiler == 'amdclang++' && inputs.platform == 'ubuntu' }} + AIS_USE_CODE_COVERAGE: ${{ inputs.cxx_compiler == 'amdclang++' && inputs.platform == 'ubuntu' && inputs.generate_coverage }} AIS_HIP_ARCHITECTURES: gfx950;gfx1201;gfx1200;gfx1101;gfx1100;gfx1030;gfx942;gfx90a;gfx908 on: workflow_call: @@ -34,6 +35,15 @@ on: cxx_compiler: required: true type: string + cxx_standard: + required: false + type: number + default: 17 + generate_coverage: + default: true + description: "Allow coverage to be generated" + required: false + type: boolean job_designator: description: "Qualifies the type of job building hipFile." required: false @@ -67,7 +77,7 @@ permissions: packages: read jobs: compile_on_AMD: - name: compile_on_AMD (${{ inputs.cxx_compiler }}) + name: compile_on_AMD (${{ inputs.cxx_compiler }} C++${{ inputs.cxx_standard}}) outputs: ais_hipfile_pkg_dev_filename: ${{ steps.pkg-metadata.outputs.AIS_HIPFILE_PKG_DEV_FILENAME }} ais_hipfile_pkg_filename: ${{ steps.pkg-metadata.outputs.AIS_HIPFILE_PKG_FILENAME }} @@ -90,7 +100,7 @@ jobs: # We should expect that there are multiple instances of this job with different cxx_compilers. # On non-pull_request triggering workflows, AIS_PR_NUMBER may be empty. run: | - echo "AIS_CONTAINER_NAME=${AIS_PR_NUMBER:=${GITHUB_RUN_ID}}_${GITHUB_JOB}_${AIS_INPUT_PLATFORM}_${AIS_INPUT_ROCM_VERSION}_${AIS_SAFE_COMPILER_NAME}" >> "${GITHUB_ENV}" + echo "AIS_CONTAINER_NAME=${AIS_PR_NUMBER:=${GITHUB_RUN_ID}}_${GITHUB_JOB}_${AIS_INPUT_PLATFORM}_${AIS_INPUT_ROCM_VERSION}_${AIS_SAFE_COMPILER_NAME}_${AIS_INPUT_CXX_STANDARD}" >> "${GITHUB_ENV}" - name: Fetching code repository... uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd #v6.0.2 with: @@ -142,6 +152,7 @@ jobs: cmake \ -DCMAKE_BUILD_TYPE=Debug \ -DCMAKE_CXX_COMPILER="${_AIS_INPUT_CXX_COMPILER}" \ + -DAIS_CXX_STANDARD="${{ env.AIS_INPUT_CXX_STANDARD }}" \ -DCMAKE_CXX_FLAGS="-Werror" \ -DCMAKE_HIP_ARCHITECTURES="${{ env.AIS_HIP_ARCHITECTURES }}" \ -DCMAKE_HIP_PLATFORM=amd \