From fdbafd7757add4509c1dc3489f29e3dddf2fde5f Mon Sep 17 00:00:00 2001 From: Scott Todd Date: Tue, 4 Nov 2025 14:29:58 -0800 Subject: [PATCH] Revert "SWDEV-554372 - Add 3 HIP_GET_PROC_ADDRESS_xxx flags (#1057)" (#1690) Reverts ROCm/rocm-systems#1057 Suspected of breaking the build, see https://github.com/ROCm/rocm-systems/pull/1057#issuecomment-3487715129 Logs: https://github.com/ROCm/rocm-systems/actions/runs/19062134668/job/54444052479#step:12:315 ``` [rocprofiler-sdk] FAILED: source/lib/rocprofiler-sdk/CMakeFiles/rocprofiler-sdk-object-library.dir/hip/abi.cpp.o [rocprofiler-sdk] ccache /opt/rh/gcc-toolset-12/root/usr/bin/c++ -DAMD_INTERNAL_BUILD=1 -DGLOG_USE_GLOG_EXPORT -DROCPROFILER_DL=1 -DROCPROFILER_HAS_GHC_LIB_FILESYSTEM=1 -DROCPROFILER_SDK_USE_SYSTEM_RCCL=0 -DROCPROFILER_SDK_USE_SYSTEM_ROCDECODE=0 -DROCPROFILER_SDK_USE_SYSTEM_ROCJPEG=0 -DUSE_PROF_API=1 -DYAML_CPP_STATIC_DEFINE -D__HIP_PLATFORM_AMD__=1 -Drocprofiler_EXPORTS=1 -I/__w/rocm-systems/rocm-systems/TheRock/build/profiler/rocprofiler-sdk/build/source/include -I/__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/source/include -I/__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/source -I/__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/yaml-cpp/include -I/__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/ptl/source -I/__w/rocm-systems/rocm-systems/TheRock/build/profiler/rocprofiler-sdk/build/external/ptl/source -isystem /__w/rocm-systems/rocm-systems/TheRock/build/core/clr/dist/include -isystem /__w/rocm-systems/rocm-systems/TheRock/build/core/ROCR-Runtime/dist/include -isystem /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/filesystem/include -isystem /__w/rocm-systems/rocm-systems/TheRock/build/profiler/rocprofiler-sdk/build/external/glog -isystem /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/glog/src -isystem /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/fmt/include -isystem /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/external/elfio -isystem /__w/rocm-systems/rocm-systems/TheRock/build/compiler/amd-comgr-stub/dist/include -isystem /__w/rocm-systems/rocm-systems/TheRock/build/third-party/sysdeps/linux/libdrm/build/stage/lib/rocm_sysdeps/lib/pkgconfig/../../include -isystem /__w/rocm-systems/rocm-systems/TheRock/build/third-party/sysdeps/linux/libdrm/build/stage/lib/rocm_sysdeps/lib/pkgconfig/../../include/libdrm -isystem /__w/rocm-systems/rocm-systems/TheRock/build/third-party/sysdeps/linux/elfutils/build/dist/lib/rocm_sysdeps/include -O3 -DNDEBUG -std=c++17 -fPIC -fvisibility=hidden -fvisibility-inlines-hidden -W -Wall -Wno-unknown-pragmas -faligned-new -rdynamic -fstack-protector-strong -Wstack-protector -MD -MT source/lib/rocprofiler-sdk/CMakeFiles/rocprofiler-sdk-object-library.dir/hip/abi.cpp.o -MF source/lib/rocprofiler-sdk/CMakeFiles/rocprofiler-sdk-object-library.dir/hip/abi.cpp.o.d -o source/lib/rocprofiler-sdk/CMakeFiles/rocprofiler-sdk-object-library.dir/hip/abi.cpp.o -c /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/source/lib/rocprofiler-sdk/hip/abi.cpp [rocprofiler-sdk] In file included from /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/source/lib/rocprofiler-sdk/hip/abi.cpp:26: [rocprofiler-sdk] /__w/rocm-systems/rocm-systems/projects/rocprofiler-sdk/source/lib/common/abi.hpp:62:27: error: static assertion failed: size of the API table struct has changed. Update the STEP_VERSION number (or in rare cases, the MAJOR_VERSION number) [rocprofiler-sdk] 62 | sizeof(TABLE) == ::rocprofiler::common::abi::compute_table_offset(NUM), \ ``` --- .../hip/amd_detail/amd_hip_runtime_pt_api.h | 3 -- .../include/hip/amd_detail/hip_api_trace.hpp | 3 -- projects/clr/hipamd/src/amdhip.def | 1 - projects/clr/hipamd/src/hip_api_trace.cpp | 6 +-- projects/clr/hipamd/src/hip_device.cpp | 54 +++++-------------- projects/clr/hipamd/src/hip_hcc.map.in | 1 - .../clr/hipamd/src/hip_table_interface.cpp | 6 --- projects/hip/include/hip/hip_runtime_api.h | 13 ----- 8 files changed, 14 insertions(+), 73 deletions(-) diff --git a/projects/clr/hipamd/include/hip/amd_detail/amd_hip_runtime_pt_api.h b/projects/clr/hipamd/include/hip/amd_detail/amd_hip_runtime_pt_api.h index 0c138c6c80..e477ad8b67 100644 --- a/projects/clr/hipamd/include/hip/amd_detail/amd_hip_runtime_pt_api.h +++ b/projects/clr/hipamd/include/hip/amd_detail/amd_hip_runtime_pt_api.h @@ -85,7 +85,6 @@ THE SOFTWARE. // Driver Entry Point API #define hipGetDriverEntryPoint __HIP_API_SPT(hipGetDriverEntryPoint) -#define hipGetProcAddress __HIP_API_SPT(hipGetProcAddress) #endif #ifdef __cplusplus @@ -198,8 +197,6 @@ hipError_t hipStreamGetCaptureInfo_v2_spt(hipStream_t stream, hipError_t hipLaunchHostFunc_spt(hipStream_t stream, hipHostFn_t fn, void* userData); hipError_t hipGetDriverEntryPoint_spt(const char* symbol, void** funcPtr, unsigned long long flags, hipDriverEntryPointQueryResult* status); -hipError_t hipGetProcAddress_spt(const char* symbol, void** pfn, int hipVersion, uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus); #ifdef __cplusplus diff --git a/projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp b/projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp index 9941eb8a79..80209d7a9c 100644 --- a/projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp +++ b/projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp @@ -1111,8 +1111,6 @@ typedef hipError_t (*t_hipLibraryEnumerateKernels)(hipKernel_t* kernels, unsigne hipLibrary_t library); typedef hipError_t (*t_hipKernelGetLibrary)(hipLibrary_t* library, hipKernel_t kernel); typedef hipError_t (*t_hipKernelGetName)(const char** name, hipKernel_t kernel); -typedef hipError_t (*t_hipGetProcAddress_spt)(const char* symbol, void** pfn, int hipVersion, uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus); // HIP Compiler dispatch table struct HipCompilerDispatchTable { @@ -1698,7 +1696,6 @@ struct HipDispatchTable { // HIP_RUNTIME_API_TABLE_STEP_VERSION == 18 t_hipOccupancyAvailableDynamicSMemPerBlock hipOccupancyAvailableDynamicSMemPerBlock_fn; - t_hipGetProcAddress_spt hipGetProcAddress_spt_fn; // DO NOT EDIT ABOVE! // HIP_RUNTIME_API_TABLE_STEP_VERSION == 19 diff --git a/projects/clr/hipamd/src/amdhip.def b/projects/clr/hipamd/src/amdhip.def index 6965222310..7407d2d7bd 100644 --- a/projects/clr/hipamd/src/amdhip.def +++ b/projects/clr/hipamd/src/amdhip.def @@ -521,4 +521,3 @@ hipLibraryEnumerateKernels hipKernelGetLibrary hipKernelGetName hipOccupancyAvailableDynamicSMemPerBlock -hipGetProcAddress_spt diff --git a/projects/clr/hipamd/src/hip_api_trace.cpp b/projects/clr/hipamd/src/hip_api_trace.cpp index 9b495f573b..a4d73f364b 100644 --- a/projects/clr/hipamd/src/hip_api_trace.cpp +++ b/projects/clr/hipamd/src/hip_api_trace.cpp @@ -790,8 +790,6 @@ hipError_t hipTexRefGetBorderColor(float* pBorderColor, const textureReference* hipError_t hipTexRefGetArray(hipArray_t* pArray, const textureReference* texRef); hipError_t hipGetProcAddress(const char* symbol, void** pfn, int hipVersion, uint64_t flags, hipDriverProcAddressQueryResult* symbolStatus = NULL); -hipError_t hipGetProcAddress_spt(const char* symbol, void** pfn, int hipVersion, uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus = NULL); hipError_t hipStreamBeginCaptureToGraph(hipStream_t stream, hipGraph_t graph, const hipGraphNode_t* dependencies, const hipGraphEdgeData* dependencyData, @@ -1379,7 +1377,6 @@ void UpdateDispatchTable(HipDispatchTable* ptrDispatchTable) { ptrDispatchTable->hipTexRefGetBorderColor_fn = hip::hipTexRefGetBorderColor; ptrDispatchTable->hipTexRefGetArray_fn = hip::hipTexRefGetArray; ptrDispatchTable->hipGetProcAddress_fn = hip::hipGetProcAddress; - ptrDispatchTable->hipGetProcAddress_spt_fn = hip::hipGetProcAddress_spt; ptrDispatchTable->hipStreamBeginCaptureToGraph_fn = hip::hipStreamBeginCaptureToGraph; ptrDispatchTable->hipGetFuncBySymbol_fn = hip::hipGetFuncBySymbol; ptrDispatchTable->hipSetValidDevices_fn = hip::hipSetValidDevices; @@ -2107,14 +2104,13 @@ HIP_ENFORCE_ABI(HipDispatchTable, hipKernelGetLibrary_fn, 503); HIP_ENFORCE_ABI(HipDispatchTable, hipKernelGetName_fn, 504); // HIP_RUNTIME_API_TABLE_STEP_VERSION == 18 HIP_ENFORCE_ABI(HipDispatchTable, hipOccupancyAvailableDynamicSMemPerBlock_fn, 505); -HIP_ENFORCE_ABI(HipDispatchTable, hipGetProcAddress_spt_fn, 506); // if HIP_ENFORCE_ABI entries are added for each new function pointer in the table, the number below // will be +1 of the number in the last HIP_ENFORCE_ABI line. E.g.: // // HIP_ENFORCE_ABI(, , 8) // // HIP_ENFORCE_ABI_VERSIONING(
, 9) <- 8 + 1 = 9 -HIP_ENFORCE_ABI_VERSIONING(HipDispatchTable, 507) +HIP_ENFORCE_ABI_VERSIONING(HipDispatchTable, 506) static_assert(HIP_RUNTIME_API_TABLE_MAJOR_VERSION == 0 && HIP_RUNTIME_API_TABLE_STEP_VERSION == 18, "If you get this error, add new HIP_ENFORCE_ABI(...) code for the new function " diff --git a/projects/clr/hipamd/src/hip_device.cpp b/projects/clr/hipamd/src/hip_device.cpp index 2869822c14..07708033fc 100644 --- a/projects/clr/hipamd/src/hip_device.cpp +++ b/projects/clr/hipamd/src/hip_device.cpp @@ -766,83 +766,55 @@ hipError_t hipGetDevicePropertiesR0000(hipDeviceProp_tR0000* prop, int device) { HIP_RETURN(hipSuccess); } -hipError_t hipGetProcAddress_common(const char* symbol, void** pfn, int hipVersion, uint64_t flags, +hipError_t hipGetProcAddress(const char* symbol, void** pfn, int hipVersion, uint64_t flags, hipDriverProcAddressQueryResult* symbolStatus) { - if (symbol == nullptr || std::string_view{symbol}.empty() || pfn == nullptr) { - return hipErrorInvalidValue; - } + HIP_INIT_API(hipGetProcAddress, symbol, pfn, hipVersion, flags, symbolStatus); + std::string symbolString = symbol; - if (flags != HIP_GET_PROC_ADDRESS_DEFAULT && flags != HIP_GET_PROC_ADDRESS_LEGACY_STREAM - && flags != HIP_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM) { - return hipErrorInvalidValue; + if (symbol == nullptr || symbolString == "" || pfn == nullptr) { + HIP_RETURN(hipErrorInvalidValue); } - bool checkSpt = (flags == HIP_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM); if (symbolString == "hipGetDeviceProperties") { if (hipVersion >= 600) { symbolString = "hipGetDevicePropertiesR0600"; } - checkSpt = false; } else if (symbolString == "hipChooseDevice") { if (hipVersion >= 600) { symbolString = "hipChooseDeviceR0600"; } - checkSpt = false; } else if (symbolString == "hipAmdFileRead") { *pfn = reinterpret_cast(&hipAmdFileRead); if (symbolStatus != nullptr) { *symbolStatus = HIP_GET_PROC_ADDRESS_SUCCESS; } - return hipSuccess; + HIP_RETURN(hipSuccess); } else if (symbolString == "hipAmdFileWrite") { *pfn = reinterpret_cast(&hipAmdFileWrite); if (symbolStatus != nullptr) { *symbolStatus = HIP_GET_PROC_ADDRESS_SUCCESS; } - return hipSuccess; + HIP_RETURN(hipSuccess); } void* handle = hip::PlatformState::instance().getDynamicLibraryHandle(); if (handle == nullptr) { - return hipErrorInvalidValue; - } - - if (checkSpt) { - symbolString += "_spt"; + HIP_RETURN(hipErrorInvalidValue); } *pfn = amd::Os::getSymbol(handle, symbolString.c_str()); - if (*pfn == nullptr) { - if (checkSpt) { - *pfn = amd::Os::getSymbol(handle, symbol); - } - if (*pfn == nullptr) { - if (symbolStatus != nullptr) { - *symbolStatus = HIP_GET_PROC_ADDRESS_SYMBOL_NOT_FOUND; - } - return hipErrorInvalidValue; + if (!(*pfn)) { + if (symbolStatus != nullptr) { + *symbolStatus = HIP_GET_PROC_ADDRESS_SYMBOL_NOT_FOUND; } + HIP_RETURN(hipErrorInvalidValue); } if (symbolStatus != nullptr) { *symbolStatus = HIP_GET_PROC_ADDRESS_SUCCESS; } - - return hipSuccess; -} - -hipError_t hipGetProcAddress(const char* symbol, void** pfn, int hipVersion, uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus) { - HIP_INIT_API(hipGetProcAddress, symbol, pfn, hipVersion, flags, symbolStatus); - HIP_RETURN(hipGetProcAddress_common(symbol, pfn, hipVersion, flags, symbolStatus)); -} - -hipError_t hipGetProcAddress_spt(const char* symbol, void** pfn, int hipVersion, uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus) { - HIP_INIT_API(hipGetProcAddress, symbol, pfn, hipVersion, flags, symbolStatus); - flags = (flags == HIP_GET_PROC_ADDRESS_DEFAULT) ? HIP_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM : flags; - HIP_RETURN(hipGetProcAddress_common(symbol, pfn, hipVersion, flags, symbolStatus)); + HIP_RETURN(hipSuccess); } } // namespace hip diff --git a/projects/clr/hipamd/src/hip_hcc.map.in b/projects/clr/hipamd/src/hip_hcc.map.in index fc915143e8..f26fc61be1 100644 --- a/projects/clr/hipamd/src/hip_hcc.map.in +++ b/projects/clr/hipamd/src/hip_hcc.map.in @@ -643,7 +643,6 @@ global: hipKernelGetLibrary; hipKernelGetName; hipOccupancyAvailableDynamicSMemPerBlock; - hipGetProcAddress_spt; local: *; } hip_7.1; diff --git a/projects/clr/hipamd/src/hip_table_interface.cpp b/projects/clr/hipamd/src/hip_table_interface.cpp index 3e378fb4d2..2165924089 100644 --- a/projects/clr/hipamd/src/hip_table_interface.cpp +++ b/projects/clr/hipamd/src/hip_table_interface.cpp @@ -1811,12 +1811,6 @@ extern "C" hipError_t hipGetProcAddress(const char* symbol, void** pfn, int hipV return hip::GetHipDispatchTable()->hipGetProcAddress_fn(symbol, pfn, hipVersion, flags, symbolStatus); } -extern "C" hipError_t hipGetProcAddress_spt(const char* symbol, void** pfn, int hipVersion, - uint64_t flags, - hipDriverProcAddressQueryResult* symbolStatus) { - return hip::GetHipDispatchTable()->hipGetProcAddress_spt_fn(symbol, pfn, hipVersion, flags, - symbolStatus); -} hipError_t hipStreamBeginCaptureToGraph(hipStream_t stream, hipGraph_t graph, const hipGraphNode_t* dependencies, const hipGraphEdgeData* dependencyData, diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index 1728b22b1b..c1226c389c 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -604,19 +604,6 @@ typedef enum hipDeviceAttribute_t { // Extended attributes for vendors } hipDeviceAttribute_t; -// Flags that can be used with hipGetProcAddress. -/** Default flag. Equivalent to HIP_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM if compiled with - * -fgpu-default-stream=per-thread flag or HIP_API_PER_THREAD_DEFAULT_STREAM macro is - * defined.*/ -#define HIP_GET_PROC_ADDRESS_DEFAULT 0x0 - -/** Search for all symbols except the corresponding per-thread versions.*/ -#define HIP_GET_PROC_ADDRESS_LEGACY_STREAM 0x1 - -/** Search for all symbols including the per-thread versions. If a per-thread version cannot be - * found, returns the legacy version.*/ -#define HIP_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM 0x2 - typedef enum hipDriverProcAddressQueryResult { HIP_GET_PROC_ADDRESS_SUCCESS = 0, HIP_GET_PROC_ADDRESS_SYMBOL_NOT_FOUND = 1,