[Sandbox] Move the Fortran bindings to one module per library (ROCm only) - #540
Open
amontoison wants to merge 5 commits into
Open
amontoison wants to merge 5 commits into
amontoison wants to merge 5 commits into
Conversation
amontoison
force-pushed
the
rocm-fortran
branch
2 times, most recently
from
September 20, 2026 16:04
7794245 to
ab3686f
Compare
amontoison
force-pushed
the
rocm-fortran
branch
2 times, most recently
from
September 20, 2026 22:55
bf7a1d1 to
fd2fee6
Compare
amontoison
force-pushed
the
rocm-fortran
branch
from
September 21, 2026 00:54
fd2fee6 to
6e02bff
Compare
Replace the single flat lib/hipfort/ with the layout the bindings would have
once they are co-located with the C headers they wrap, one directory per
library grouped by upstream repository:
lib/rocm-systems/{hip,roctx}/
lib/rocm-libraries/{rocblas,hipblas,rocsolver,hipsolver,rocsparse,hipsparse,
rocfft,hipfft,hipfftw,rocrand,hiprand}/
The module names, the generic names and the bind(C) symbols are unchanged, so
no user code and no test needs to be touched. Each library now ships its module
plus two included fragments (<mod>_interfaces.f90, <mod>_wrappers.f90), which
keeps a single library's surface reviewable on its own.
lib/CMakeLists.txt changes accordingly:
* the source list is built per directory instead of from one flat glob, and
excludes *_interfaces.f90 / *_wrappers.f90. Those are INCLUDEd by their
module, not compilation units: compiling them standalone fails, since they
are bare interface and contains bodies with no enclosing module.
* every library directory is added to the include path. gfortran resolves an
INCLUDE relative to the including file, but that is not guaranteed by the
standard, so the directories are named explicitly for other compilers.
* hipfort_cuda_errors.f90 is hand-written rather than generated; it moves to
rocm-systems/hip/ with the rest of the HIP runtime.
Validated on gfortran with -DHIPFORT_ASSUMED_RANK=ON: 901 Fortran objects build
with no error, and ctest reports 843 of 863 passing on a gfx1101 GPU. The 20
failures are identical to those the flat bindings produce, and each was checked
to reproduce against upstream's own unmodified lib/hipfort: 16 rocSPARSE
csrgeam/csrgemm SIGSEGV inside librocsparse's nnz phase, 1 sddmm_batched
returning rocsparse_status 2 (not implemented), 2 hipMemRangeGetAttribute
ReadMostly unsupported on this device, 1 hipModuleLaunchKernel returning 719.
hipBLAS is skipped at configure time on this machine because its config package
cannot find hipblas-common, so its 79 tests are not covered by the run above.
Replace the flat lib/hipfort/ with the shape the bindings take once they are
co-located with the C headers they wrap: one self-contained module per library,
named after the library, grouped by upstream repository.
lib/rocm-systems/hip/
lib/rocm-libraries/{rocblas,hipblas,rocsolver,hipsolver,rocsparse,hipsparse,
rocfft,hipfft,hipfftw,rocrand,hiprand}/
Each module folds in what used to be scattered across sibling modules: the
enums, the derived types, the opaque handles and the <lib>Check helper. Twelve
files replace sixty-three, and a library's whole Fortran surface is one file.
The only cross-library dependency left is rocsolver naming six rocBLAS entities
in an `only` clause, which mirrors rocsolver.h including rocblas.h.
This renames the modules: hipfort_hipblas and its _enums/_types become hipblas,
the HIP runtime modules all become hip. Every test's `use` lines were rewritten
accordingly (3728 rewritten, 1543 now-duplicate uses collapsed). No procedure,
constant or bind(C) symbol name changes, so only the `use` lines move.
ROCm only. The generator loads the CUDA symbol maps solely for the flat layout,
so these sources carry no USE_CUDA_NAMES branch at all. Rather than let an
"nvptx" build silently produce an archive still bound to hip*/roc* symbols,
HIPFORT_BUILD_NVPTX now raises a fatal error, and hipfort_cuda_errors.f90 goes
with the rest of the CUDA support. Teaching the per-library layout to emit the
CUDA bindings is a separate piece of work.
rocTX is left out. It exists as two incompatible APIs: roctracer/roctx.h with
libroctx64 exports versioned symbols (roctxMarkA@@ROCTX_4.1) and declares
roctx_version_major/minor, while rocprofiler-sdk-roctx/roctx.h with
librocprofiler-sdk-roctx exports plain symbols, drops those two and adds
roctxProfilerPause/Resume, roctxGetThreadId and the roctxName* family.
hipfort::roctx links the latter while the bindings are generated from the
former. Choosing a side is an API decision of its own and does not belong in a
layout change, so no rocTX module ships here and its tests are guarded off with
their sources left in place.
Validated with gfortran, -DHIPFORT_ASSUMED_RANK=ON -DHIPFORT_EXTENDED_TESTS=ON:
every target builds, and ctest reports 855 of 875 passing on a gfx1101 GPU. The
20 failures are identical to those the flat bindings produce, and each was
checked to reproduce against upstream's own unmodified lib/hipfort: 16 rocSPARSE
csrgeam/csrgemm SIGSEGV inside librocsparse's nnz phase, 1 sddmm_batched
returning rocsparse_status 2 (not implemented), 2 hipMemRangeGetAttribute
ReadMostly unsupported on this device, 1 hipModuleLaunchKernel returning 719.
The same configuration under amdflang gives the same 20.
hipBLAS is skipped at configure time on this machine because its config package
cannot find hipblas-common; its tests were compiled and linked by hand instead
(137 of 137 clean).
The prose still told readers to `use hipfort_rocblas` and to reach for `hipfort_check`, neither of which exists in this layout. Updating it makes the proof-of-concept complete: a reader can follow the tutorials against these bindings without hitting a module that is not there. Most of it is the rename, but three groups of sentences described a structure that no longer exists and had to be rewritten rather than substituted: * The `<lib>Check` helpers are no longer in a shared `hipfort_check` module; each one travels with the library it checks. `rocblasCheck` now comes from `rocblas`, `hipsparseCheck` from `hipsparse`, and only `hipCheck` from `hip`. A blind rename would have pointed every one of them at `hip`. * The `<lib>_enums` modules are folded in, so sentences of the form "the enumerations come from `hipfort_rocblas_enums`, which `hipfort_rocblas` re-exports" degenerated into a module re-exporting itself. * The HIP page listed where each piece lived: enumerators in `hipfort_enums`, types in `hipfort_types`, `hipGetDeviceProperties` in `hipfort_auxiliary`, and so on. That whole inventory collapses to one sentence now that `hip` holds all of it. Two `hipfort_` strings are deliberately untouched: `-Dhipfort_ROOT` is a CMake variable, and `hipfort_test_f2008_hipblas_dgemm` is a ctest target name. Neither is a Fortran module. The generated supported-API tables under docs/doxygen/input/ still carry `hipfort_`-prefixed anchors. They are produced from the bindings, so they follow whenever the generator is pointed at this layout.
Three globs and two snippets still named `lib/hipfort/`, which this branch deletes. None of them fail loudly: a glob that matches nothing produces an empty file list, and an empty file list builds fine. * `docs/conf.py` preprocessed `../lib/hipfort/*.[fF]90` into `docs/doxygen/input/`, which the Doxyfile then reads. The glob now matched zero files, so the whole API reference would have been generated from no sources at all. * `test/openmp/CMakeLists.txt` compiled the bindings straight from the same path. That project is standalone, not reached by `add_subdirectory`, so the main build never noticed that `test_rocsolver_zhegvdx` had lost every module it depends on. * `README.md` still opened both of its Fortran examples with `use hipfort`, and `docs/README.md` documented the old preprocessing path. Also adds the changelog entry for the restructure itself, which had none, and stops the ROCTx entries from naming a `hipfort_roctx` module that no longer exists under that name. An empty `lib/hipfort/` may linger in a working tree that predates the move; git does not track directories, so it is only a local artefact.
amontoison
force-pushed
the
rocm-fortran
branch
from
September 21, 2026 01:24
6e02bff to
c0c5a93
Compare
The tree was generated with the convenience specifics on: native array overloads, typed handles, character(len=*) forms. That is 803k lines of Fortran, and it makes this branch hard to compare against what hipfort ships. Regenerated with NATIVE/TYPED/CSTR=0, the same flags regenerate.sh uses for the compatibility track, so each module now exposes the plain bind(C) surface and nothing else. 264k lines. rocblas.F90 alone drops from 179k to 67k. The derived types go with them, and that is not a loss: emit_types only fires when some signature references a type, and with handles as type(c_ptr) and buffers as type(c_ptr) none does. hipfort ships three *_types modules for the same reason. Adding a convenience back later is a small change; removing one from an API people already call is not. So this track starts at parity with hipfort and grows from there. The generated symbol tests are regenerated to match: without a generic, a routine is a bare bind(C) named after the C symbol rather than a <name>_ specific, so the old tests referenced names that no longer exist. Built with -DHIPFORT_EXTENDED_TESTS=ON and run on a gfx1101 GPU: 992 of 1011 tests pass, and every failure is in the environment set this branch already documents.
This was referenced Sep 21, 2026
This branch has not been deployed
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
A sandbox preview, not for merge.
It shows what the Fortran bindings look like once they are co-located with the C headers they wrap, so the shape can be discussed before anything is committed to.
The layout
lib/hipfortis a single flat directory of 33 files today.This replaces it with one directory per library, grouped by upstream repository:
Each library is now a single self-contained module.
What used to be scattered across sibling modules (the enums, the derived types, the opaque handles, the check helper) is folded in.
--> 12 files replace 33.