hipFORT

https://rocm.docs.amd.com/projects/hipfort/en/latest/
Admin
hipFORT is a Fortran interface library for accessing GPU kernels that provides standard interfaces...
Tokens:8,973,621
Snippets:30,279
Trust Score:10
Update:1 month ago
Show doc for...
Tokens:
Raw
### rocfft_setup_()

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocfft_1_1rocfft__setup.html

Library setup function, called at least once in program before start of library use.

```APIDOC
## rocfft_setup_()

### Description
Library setup function, called at least once in program before start of library use.

### Function Signature
integer(kind(rocfft_status_success)) function hipfort_rocfft::rocfft_setup::rocfft_setup_

### Parameters
None

### Return Value
integer(kind(rocfft_status_success)) - Status code indicating success or failure.

### Usage
Call this function once before using any other rocFFT routines.
```

--------------------------------

### rocfft_setup

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort__rocfft_8F90.html

Library setup function, called at least once in program before start of library use.

```APIDOC
## rocfft_setup

### Description
Library setup function, called at least once in program before start of library use.

### Interface
`interface hipfort_rocfft::rocfft_setup`

### Usage
Call this function before any other rocFFT operations.

```

--------------------------------

### Build and test hipFORT from source

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/install/quick-start.html

Standard commands to clone the repository, configure the build with testing enabled, compile, install, and run the test suite.

```bash
git clone https://github.com/ROCm/hipfort.git
cd hipfort
cmake -S. -Bbuild -DCMAKE_INSTALL_PREFIX=/tmp/hipfort -DBUILD_TESTING=ON
cmake --build build
cmake --install build
ctest --test-dir build
```

--------------------------------

### C example: rocsparse_get_version usage

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__get__version.html

Shows how to get the rocSPARSE version using the C API. It creates a handle, calls rocsparse_get_version, and destroys the handle.

```c
rocsparse_handle handle;
rocsparse_create_handle(&handle);
rocsparse_get_version(handle, &rocsparse_ver);
rocsparse_destroy_handle(handle);
```

--------------------------------

### hipfort::hiplaunchbyptr

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort_8F90.html

Launch a kernel.

```APIDOC
## hipfort::hiplaunchbyptr

### Description
Launch a kernel.

### Interface
`hipfort::hiplaunchbyptr`

### Parameters
No parameters are documented in the source.

### Returns
No return value is documented in the source.
```

--------------------------------

### Link hipFORT with CMake targets

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/install/install.html

Use this CMake snippet to link hipFORT into your project. It requires the hipfort package to be installed and provides exported targets such as hipfort::hip, hipfort::rocblas, and hipfort::hipblas. The example links against hipfort::hipblas and hipfort::hip.

```cmake
find_package(hipfort REQUIRED)
add_executable(my_app main.f08)
target_link_libraries(my_app PRIVATE hipfort::hipblas hipfort::hip)

```

--------------------------------

### Check CSC matrix validity with rocsparse_scheck_matrix_csc

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__scheck__matrix__csc.html

Checks a CSC matrix for validity using rocsparse_scheck_matrix_csc. The example uses an invalid matrix with a duplicate row index in the second column, so the data_status output will indicate an issue. It first calls rocsparse_scheck_matrix_csc_buffer_size to get the required buffer size, then allocates that buffer and passes it to rocsparse_scheck_matrix_csc. The routine does not support execution in a hipGraph context.

```cpp
// 1 2 0 0
// 0 3 4 0
// 2 0 1 1
// 0 3 0 2
std::vector<int> hcsc_row_ind = {0, 2, 0, 1, 1, 1, 2, 2, 3}; //<---duplicate row index in
second column
std::vector<int> hcsc_col_ptr = {0, 2, 5, 7, 9};
std::vector<float> hcsc_val = {1, 2, 2, 3, 3, 4, 1, 1, 2};
int m = 4;
int n = 4;
int nnz = 9;
int* dcsc_row_ind = nullptr;
int* dcsc_col_ptr = nullptr;
float* dcsc_val = nullptr;
hipMalloc((void**)&dcsc_row_ind, sizeof(int) * nnz);
hipMalloc((void**)&dcsc_col_ptr, sizeof(int) * (n + 1));
hipMalloc((void**)&dcsc_val, sizeof(float) * nnz);
hipMemcpy(dcsc_row_ind, hcsc_row_ind.data(), sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsc_col_ptr, hcsc_col_ptr.data(), sizeof(int) * (n + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsc_val, hcsc_val.data(), sizeof(float) * nnz, hipMemcpyHostToDevice);
rocsparse_handle handle;
rocsparse_create_handle(&handle);
const rocsparse_index_base idx_base = rocsparse_index_base_zero;
const rocsparse_fill_mode fill_mode = rocsparse_fill_mode_upper;
const rocsparse_matrix_type matrix_type = rocsparse_matrix_type_general;
const rocsparse_storage_mode storage_mode = rocsparse_storage_mode_sorted;
rocsparse_data_status data_status;
size_t buffer_size;
rocsparse_scheck_matrix_csc_buffer_size(handle, m, n, nnz, dcsc_val, dcsc_col_ptr,
dcsc_row_ind,
idx_base, matrix_type, fill_mode, storage_mode, &buffer_size);
void* dbuffer = nullptr;
hipMalloc((void**)&dbuffer, buffer_size);
rocsparse_scheck_matrix_csc(handle, m, n, nnz, dcsc_val, dcsc_col_ptr, dcsc_row_ind,
idx_base,
matrix_type, fill_mode, storage_mode, &data_status, dbuffer);
std::cout << "data_status: " << data_status << std::endl;
hipFree(dbuffer);
rocsparse_destroy_handle(handle);
hipFree(dcsc_row_ind);
hipFree(dcsc_col_ptr);
hipFree(dcsc_val);
```

--------------------------------

### hiplaunchbyptr_()

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hiplaunchbyptr.html

Documentation for the hiplaunchbyptr_ function, which launches a kernel.

```APIDOC
## hiplaunchbyptr_()

### Description
Launches a kernel. This function is part of the hipfort interface and is used to launch a kernel by pointer.

### Method
Function

### Signature
integer(kind(hipsuccess)) function hipfort::hiplaunchbyptr::hiplaunchbyptr_(type(c_ptr), value _func_)

### Parameters
- **func** (type(c_ptr)) - Input - Kernel to launch.

### Returns
- `hipSuccess`, `hipErrorNotInitialized`, `hipErrorInvalidValue`
```

--------------------------------

### hipblasaxpybatchedex code examples

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__hipblas_1_1hipblasaxpybatchedex.html

Code examples extracted from the documentation for the hipblasaxpybatchedex interface.

```text
y := alpha * x + y
```

```fortran
integer(kind(hipblas_status_success)) function hipblasaxpybatchedex_ (handle, n, alpha, alphatype, x, xtype, incx, y, ytype, incy, batchcount, executiontype)
```

```fortran
integer(kind(hipblas_status_success)) function hipblasaxpybatchedex_  | ( | type(c_ptr), value  | _handle_ ,   
|  | integer(c_int), value  | _n_ ,   
|  | type(c_ptr), value  | _alpha_ ,   
|  | integer(kind(hip_r_32f)), value  | _alphatype_ ,   
|  | type(c_ptr), value  | _x_ ,   
|  | integer(kind(hip_r_32f)), value  | _xtype_ ,   
|  | integer(c_int), value  | _incx_ ,   
|  | type(c_ptr), value  | _y_ ,   
|  | integer(kind(hip_r_32f)), value  | _ytype_ ,   
|  | integer(c_int), value  | _incy_ ,   
|  | integer(c_int), value  | _batchcount_ ,   
|  | integer(kind(hip_r_32f)), value  | _executiontype_  
| ) | |
```

--------------------------------

### Configure build with a toolchain file

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/install/install.html

Use a pre-defined CMake toolchain file from the cmake/toolchains/ directory to set compilers and backend paths.

```bash
cmake -S . -Bbuild -DCMAKE_TOOLCHAIN_FILE=cmake/toolchains/amdflang.cmake
```

--------------------------------

### hipfort::hiplaunchkernel

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort_8F90.html

C compliant kernel launch API.

```APIDOC
## hipfort::hiplaunchkernel

### Description
C compliant kernel launch API.

### Interface
`hipfort::hiplaunchkernel`

### Parameters
No parameters are documented in the source.

### Returns
No return value is documented in the source.
```

--------------------------------

### hipfort::hipprofilerstart

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort_8F90.html

Start recording of profiling information [Deprecated] When using this API, start the profiler with profiling disabled.

```APIDOC
## hipfort::hipprofilerstart

### Description
Start recording of profiling information [Deprecated] When using this API, start the profiler with profiling disabled.

### Interface
`hipfort::hipprofilerstart`

### Parameters
No parameters are documented in the source.

### Returns
No return value is documented in the source.
```

--------------------------------

### Packed Storage Format Examples

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocblas_1_1rocblas__chpr2__batched.html

Examples showing how the upper and lower triangular portions of a 3x3 matrix are compacted into packed storage.

```text
  if uplo == rocblas_fill_upper:
      The upper triangular portion of each Hermitian matrix A_i is supplied.
      The matrix is compacted so that AP contains the triangular portion
      column-by-column
      so that:
      AP(0) = A(0,0)
      AP(1) = A(0,1)
      AP(2) = A(1,1), etc.
      Ex: (rocblas_fill_upper; n = 3)
          (1, 0) (2, 1) (4,9)
          (2,-1) (3, 0) (5,3) ---> [(1,0),(2,1),(3,0),(4,9),(5,3),(6,0)]
          (4,-9) (5,-3) (6,0)

  if uplo == rocblas_fill_lower:
      The lower triangular portion of each Hermitian matrix A_i is supplied.
      The matrix is compacted so that AP contains the triangular portion
      column-by-column
      so that:
      AP(0) = A(0,0)
      AP(1) = A(1,0)
      AP(2) = A(2,1), etc.
      Ex: (rocblas_fill_lower; n = 3)
          (1, 0) (2, 1) (4,9)
          (2,-1) (3, 0) (5,3) --> [(1,0),(2,-1),(4,-9),(3,0),(5,-3),(6,0)]
          (4,-9) (5,-3) (6,0)
```

--------------------------------

### hiplaunchkernel_

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hiplaunchkernel.html

Documentation for the hiplaunchkernel_ function, a C-compliant kernel launch API.

```APIDOC
### Function: hiplaunchkernel_

**Signature:**
```fortran
integer(kind(hipsuccess)) function hipfort::hiplaunchkernel::hiplaunchkernel_(
    type(c_ptr), value :: _function_address_,
    type(c_ptr), value :: _numblocks_,
    type(c_ptr), value :: _dimblocks_,
    type(c_ptr) :: _args_,
    integer(c_size_t), value :: _sharedmembytes_,
    type(c_ptr), value :: _stream_
)
```

**Description:**
C compliant kernel launch API.

**Parameters:**
- `_function_address_` (type(c_ptr), value) - Kernel stub function pointer.
- `_numblocks_` (type(c_ptr), value) - Number of blocks.
- `_dimblocks_` (type(c_ptr), value) - Dimension of a block.
- `_args_` (type(c_ptr)) - Pointer of arguments passed to the kernel. If the kernel has multiple parameters, 'args' should be array of pointers, each points the corresponding argument.
- `_sharedmembytes_` (integer(c_size_t), value) - Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
- `_stream_` (type(c_ptr), value) - Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

**Returns:**
- `hipSuccess`, `hipErrorInvalidValue`
```

--------------------------------

### Doxygen Graph Legend Example: C++ Class Definitions

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/graph_legend.html

Example C++ class definitions used to illustrate Doxygen-generated inheritance and containment graphs. The example includes documented, undocumented, truncated, template, and inheritance classes, and results in a graph where boxes and arrows have specific meanings.

```cpp
/*! Invisible class because of truncation */
class Invisible { };
/*! Truncated class, inheritance relation is hidden */
class Truncated : public Invisible { };
/* Class not documented with doxygen comments */
class Undocumented { };
/*! Class that is inherited using public inheritance */
class PublicBase : public Truncated { };
/*! A template class */
template<class T> class Templ { };
/*! Class that is inherited using protected inheritance */
class ProtectedBase { };
/*! Class that is inherited using private inheritance */
class PrivateBase { };
/*! Class that is used by the Inherited class */
class Used { };
/*! Super class that inherits a number of other classes */
class Inherited : public PublicBase,
protected ProtectedBase,
private PrivateBase,
public Undocumented,
public Templ<int>
{
private:
Used *m_usedClass;
};
```

--------------------------------

### rocfft_plan_description_create

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort__rocfft_8F90.html

Create plan description.

```APIDOC
## rocfft_plan_description_create

### Description
Create plan description.

### Interface
`interface hipfort_rocfft::rocfft_plan_description_create`

### Usage
Use this function to create a plan description object.

```

--------------------------------

### hipmodulelaunchkernel_

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hipmodulelaunchkernel.html

Launches a kernel f with launch parameters and shared memory on a stream with arguments passed to kernelparams or extra.

```APIDOC
## hipmodulelaunchkernel_

### Description
Launches kernel f with launch parameters and shared memory on stream with arguments passed to kernelparams or extra.

### Function Signature
integer(kind(hipsuccess)) function hipmodulelaunchkernel_ (f, griddimx, griddimy, griddimz, blockdimx, blockdimy, blockdimz, sharedmembytes, stream, kernelparams, extra)

### Parameters
- **f** (in) - Kernel to launch.
- **gridDimX** (in) - X grid dimension specified as multiple of blockDimX.
- **gridDimY** (in) - Y grid dimension specified as multiple of blockDimY.
- **gridDimZ** (in) - Z grid dimension specified as multiple of blockDimZ.
- **blockDimX** (in) - X block dimensions specified in work-items.
- **blockDimY** (in) - Y grid dimension specified in work-items.
- **blockDimZ** (in) - Z grid dimension specified in work-items.
- **sharedMemBytes** (in) - Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
- **stream** (in) - Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.
- **kernelParams** (in) - Kernel parameters to launch.
- **extra** (in) - Pointer to kernel arguments. These are passed directly to the kernel and must be in the memory layout and alignment expected by the kernel. All passed arguments must be naturally aligned according to their type. The memory address of each argument should be a multiple of its size in bytes. Please refer to hip_porting_driver_api.md for sample usage.

### Notes
- HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32. So gridDim.x * blockDim.x, gridDim.y * blockDim.y and gridDim.z * blockDim.z are always less than 2^32.

### Returns
- `hipSuccess`, `hipErrorNotInitialized`, `hipErrorInvalidValue`
```

--------------------------------

### Usage Example for rocsparse_set_stream

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__set__stream.html

This example demonstrates the workflow of creating a rocSPARSE handle and a HIP stream, associating them using rocsparse_set_stream, and performing cleanup.

```cpp
// Create rocSPARSE handle
rocsparse_handle handle;
rocsparse_create_handle(&handle);
// Create stream
hipStream_t stream;
hipStreamCreate(&stream);
// Set stream to rocSPARSE handle
rocsparse_set_stream(handle, stream);
// Do some work
// ...
// Clean up
rocsparse_destroy_handle(handle);
hipStreamDestroy(stream);
```

--------------------------------

### CSRGEMM Example with rocsparse

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__scsrgemm__numeric.html

This example demonstrates the full workflow for multiplying two CSR matrices with a scalar alpha and adding the result to another CSR matrix.

```cpp
// Initialize scalar multipliers
float alpha = 2.0f;
float beta = 1.0f;
// Create matrix descriptors
rocsparse_mat_descr descr_A;
rocsparse_mat_descr descr_B;
rocsparse_mat_descr descr_C;
rocsparse_mat_descr descr_D;
rocsparse_create_mat_descr(&descr_A);
rocsparse_create_mat_descr(&descr_B);
rocsparse_create_mat_descr(&descr_C);
rocsparse_create_mat_descr(&descr_D);
// Create matrix info structure
rocsparse_mat_info info_C;
rocsparse_create_mat_info(&info_C);
// Set pointer mode
rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_host);
// Query rocsparse for the required buffer size
size_t buffer_size;
rocsparse_scsrgemm_buffer_size(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
&alpha,
descr_A,
nnz_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_row_ptr_B,
csr_col_ind_B,
&beta,
descr_D,
nnz_D,
csr_row_ptr_D,
csr_col_ind_D,
info_C,
&buffer_size);
// Allocate buffer
void* buffer;
hipMalloc(&buffer, buffer_size);
// Obtain number of total non-zero entries in C and row pointers of C
rocsparse_int nnz_C;
hipMalloc((void**)&csr_row_ptr_C, sizeof(rocsparse_int) * (m + 1));
rocsparse_csrgemm_nnz(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
descr_A,
nnz_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_row_ptr_B,
csr_col_ind_B,
descr_D,
nnz_D,
csr_row_ptr_D,
csr_col_ind_D,
descr_C,
csr_row_ptr_C,
&nnz_C,
info_C,
buffer);
// Compute column indices and values of C
hipMalloc((void**)&csr_col_ind_C, sizeof(rocsparse_int) * nnz_C);
rocsparse_csrgemm_symbolic(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
descr_A,
nnz_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_row_ptr_B,
csr_col_ind_B,
descr_D,
nnz_D,
csr_row_ptr_D,
csr_col_ind_D,
descr_C,
nnz_C,
csr_row_ptr_C,
csr_col_ind_C,
info_C,
buffer);
hipMalloc((void**)&csr_val_C, sizeof(float) * nnz_C);
rocsparse_scsrgemm_numeric(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
&alpha,
descr_A,
nnz_A,
csr_val_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_val_B,
csr_row_ptr_B,
csr_col_ind_B,
&beta,
descr_D,
nnz_D,
csr_val_D,
csr_row_ptr_D,
csr_col_ind_D,
descr_C,
nnz_C,
csr_val_C,
csr_row_ptr_C,
csr_col_ind_C,
info_C,
buffer);
```

--------------------------------

### rocfft_plan_create

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort__rocfft_8F90.html

Create an FFT plan.

```APIDOC
## rocfft_plan_create

### Description
Create an FFT plan.

### Interface
`interface hipfort_rocfft::rocfft_plan_create`

### Usage
Use this function to create a plan for FFT operations.

```

--------------------------------

### rocsparse_scoomv example

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__scoomv.html

Performs a sparse matrix vector multiplication in COO format. The example shows the loop structure for the operation, including the initialization of y and the accumulation of the product.

```c
for(i = 0; i < m; ++i)
{
y[i] = beta * y[i];
}
for(i = 0; i < nnz; ++i)
{
y[coo_row_ind[i]] += alpha * coo_val[i] * x[coo_col_ind[i]];
}
```

--------------------------------

### hipfort::hipconfigurecall

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort_8F90.html

Configures a kernel launch.

```APIDOC
## hipfort::hipconfigurecall

### Description
Configures a kernel launch.

### Interface
`hipfort::hipconfigurecall`

### Parameters
No parameters are documented in the source.

### Returns
No return value is documented in the source.
```

--------------------------------

### CSR Matrix Multiplication Example

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocsparse_1_1rocsparse__scsrgemm.html

This example demonstrates the workflow for performing sparse matrix-matrix multiplication, including buffer size calculation, non-zero entry estimation, and the final computation.

```C
// Initialize scalar multipliers
float alpha = 2.0f;
float beta = 1.0f;
// Create matrix descriptors
rocsparse_mat_descr descr_A;
rocsparse_mat_descr descr_B;
rocsparse_mat_descr descr_C;
rocsparse_mat_descr descr_D;
rocsparse_create_mat_descr(&descr_A);
rocsparse_create_mat_descr(&descr_B);
rocsparse_create_mat_descr(&descr_C);
rocsparse_create_mat_descr(&descr_D);
// Create matrix info structure
rocsparse_mat_info info_C;
rocsparse_create_mat_info(&info_C);
// Set pointer mode
rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_host);
// Query rocsparse for the required buffer size
size_t buffer_size;
rocsparse_scsrgemm_buffer_size(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
&alpha,
descr_A,
nnz_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_row_ptr_B,
csr_col_ind_B,
&beta,
descr_D,
nnz_D,
csr_row_ptr_D,
csr_col_ind_D,
info_C,
&buffer_size);
// Allocate buffer
void* buffer;
hipMalloc(&buffer, buffer_size);
// Obtain number of total non-zero entries in C and row pointers of C
rocsparse_int nnz_C;
hipMalloc((void**)&csr_row_ptr_C, sizeof(rocsparse_int) * (m + 1));
rocsparse_csrgemm_nnz(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
descr_A,
nnz_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_row_ptr_B,
csr_col_ind_B,
descr_D,
nnz_D,
csr_row_ptr_D,
csr_col_ind_D,
descr_C,
csr_row_ptr_C,
&nnz_C,
info_C,
buffer);
// Compute column indices and values of C
hipMalloc((void**)&csr_col_ind_C, sizeof(rocsparse_int) * nnz_C);
hipMalloc((void**)&csr_val_C, sizeof(float) * nnz_C);
rocsparse_scsrgemm(handle,
rocsparse_operation_none,
rocsparse_operation_none,
m,
n,
k,
&alpha,
descr_A,
nnz_A,
csr_val_A,
csr_row_ptr_A,
csr_col_ind_A,
descr_B,
nnz_B,
csr_val_B,
csr_row_ptr_B,
csr_col_ind_B,
&beta,
descr_D,
nnz_D,
csr_val_D,
csr_row_ptr_D,
csr_col_ind_D,
descr_C,
csr_val_C,
csr_row_ptr_C,
csr_col_ind_C,
info_C,
buffer);
```

--------------------------------

### hipextlaunchkernel_

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hipextlaunchkernel.html

Launches a kernel from the pointer address, with arguments and shared memory on stream.

```APIDOC
## hipextlaunchkernel_

### Description
Launches kernel from the pointer address, with arguments and shared memory on stream.

### Parameters
- **function_address** (in) - Pointer to the Kernel to launch.
- **numBlocks** (in) - Number of blocks.
- **dimBlocks** (in) - Dimension of a block.
- **args** (in) - Pointer of arguments passed to the kernel. If the kernel has multiple parameters, 'args' should be array of pointers, each points the corresponding argument.
- **sharedMemBytes** (in) - Amount of dynamic shared memory to allocate for this kernel. HIP-Clang compiler provides support for extern shared declarations.
- **stream** (in) - Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.
- **startEvent** (in) - If non-null, specified event will be updated to track the start time of the kernel launch. The event must be created before calling this API.
- **stopEvent** (in) - If non-null, specified event will be updated to track the stop time of the kernel launch. The event must be created before calling this API.
- **flags** (in) - The value of hipExtAnyOrderLaunch, signifies if kernel can be launched in any order.

### Returns
`hipSuccess`, `hipErrorNotInitialized`, `hipErrorInvalidValue`.
```

--------------------------------

### hipmemptrgetinfo_ (ptr, mysize)

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hipmemptrgetinfo.html

Get allocated memory size via memory pointer. This function gets the allocated shared virtual memory size from the given memory pointer.

```APIDOC
## hipmemptrgetinfo_

### Description
Gets the allocated shared virtual memory size from a memory pointer.

### Function

`integer(kind(hipsuccess)) function hipmemptrgetinfo_(ptr, mysize)`

### Parameters

- **ptr** (type(c_ptr), value) - [in] Pointer to allocated memory.
- **mysize** (integer(c_size_t)) - [out] Returns the allocated memory size in bytes.

### Return Value

- `hipSuccess` on success.
- `hipErrorInvalidValue` on failure.

### Interface

```fortran
integer(kind(hipsuccess)) function hipmemptrgetinfo_(type(c_ptr), value ptr, integer(c_size_t) mysize)
```

### Example
*(No example provided in the source).*
```

--------------------------------

### fftwf_plan_guru64_dft_r2c_

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__hipfftw_1_1fftwf__plan__guru64__dft__r2c.html

Creates an FFTW plan for a real-to-complex DFT using 64-bit iodim layout descriptors. Returns a c_ptr plan handle.

```APIDOC
## fftwf_plan_guru64_dft_r2c_

### Description
Creates an FFTW plan for a real-to-complex DFT using 64-bit iodim layout descriptors. Equivalent to fftwf_plan_guru_dft_r2c but uses fftwf_iodim64 values.

### Function Signature
```fortran
type(c_ptr) function fftwf_plan_guru64_dft_r2c_ (rank, dims, howmany_rank, howmany_dims, in, out, flags)
```

### Parameters
- **rank** (integer(c_int), value) - Required - Rank of the transform.
- **dims** (type(c_ptr), value) - Required - Pointer to an array of fftwf_iodim64 describing the transform dimensions.
- **howmany_rank** (integer(c_int), value) - Required - Rank of the howmany dimensions.
- **howmany_dims** (type(c_ptr), value) - Required - Pointer to an array of fftwf_iodim64 describing the howmany dimensions.
- **in** (type(c_ptr), value) - Required - Pointer to the input array.
- **out** (type(c_ptr), value) - Required - Pointer to the output array.
- **flags** (integer(c_int), value) - Required - FFTW planner flags.

### Return Value
- **type(c_ptr)** - A pointer to the created FFTW plan.

```

--------------------------------

### rocblas_fill_lower packed layout example

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocblas_1_1rocblas__sspr2__batched.html

Shows how the packed array AP is laid out when uplo is rocblas_fill_lower. The lower triangular portion of each symmetric matrix is stored column-by-column, with an example for n = 4.

```text
if uplo == rocblas_fill_lower:
    The lower triangular portion of each symmetric matrix A_i is supplied.
    The matrix is compacted so that AP contains the triangular portion
    column-by-column
    so that:
    AP(0) = A(0,0)
    AP(1) = A(1,0)
    AP(n) = A(2,1), etc.
        Ex: (rocblas_fill_lower; n = 4)
            1 2 3 4
            2 5 6 7    -----> [1, 2, 3, 4, 5, 6, 7, 8, 9, 0]
            3 6 8 9
            4 7 9 0
```

--------------------------------

### hipfftmakeplan2d

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__hipfft_1_1hipfftmakeplan2d.html

Initializes a new two-dimensional FFT plan.

```APIDOC
## hipfftmakeplan2d

### Description
Initializes a new two-dimensional FFT plan.

### Interface
```fortran
interface hipfftmakeplan2d
```

### Parameters
- **plan** (type(c_ptr)) - Pointer to the FFT plan handle.
- **nx** (integer(c_int)) - Size of the first dimension.
- **ny** (integer(c_int)) - Size of the second dimension.
- **type** (integer(c_int)) - Transform type (e.g., HIPFFT_R2C, HIPFFT_C2R, etc.).
- **batch** (integer(c_int)) - Number of transforms to execute.
- **stream** (type(c_ptr)) - Stream to use for execution.

### Return Value
- **integer(c_int)** - Status code (0 on success).

### Example
```fortran
! Example usage
integer(c_int) :: status
status = hipfftmakeplan2d(plan, nx, ny, type, batch, stream)
```
```

--------------------------------

### hipextlaunchkernel_()

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort_1_1hipextlaunchkernel.html

Launches a kernel on a HIP stream with optional start and stop events, using the provided function address, block configuration, arguments, shared memory size, and flags.

```APIDOC
## hipextlaunchkernel_()

### Description
Launches a kernel on a HIP stream with optional start and stop events, using the provided function address, block configuration, arguments, shared memory size, and flags.

### Function Signature
```fortran
integer(kind(hipsuccess)) function hipfort::hipextlaunchkernel::hipextlaunchkernel_(
  type(c_ptr), value :: _function_address_,
  type(c_ptr), value :: _numblocks_,
  type(c_ptr), value :: _dimblocks_,
  type(c_ptr) :: _args_,
  integer(c_size_t), value :: _sharedmembytes_,
  type(c_ptr), value :: _stream_,
  type(c_ptr), value :: _startevent_,
  type(c_ptr), value :: _stopevent_,
  integer(c_int), value :: _flags_
)
```

### Parameters
- **function_address** (type(c_ptr), value) - Required - Pointer to the kernel function.
- **numblocks** (type(c_ptr), value) - Required - Pointer to the number of blocks.
- **dimblocks** (type(c_ptr), value) - Required - Pointer to the block dimensions.
- **args** (type(c_ptr)) - Required - Pointer to the kernel arguments.
- **sharedmembytes** (integer(c_size_t), value) - Required - Size of shared memory in bytes.
- **stream** (type(c_ptr), value) - Required - HIP stream handle.
- **startevent** (type(c_ptr), value) - Required - Start event handle.
- **stopevent** (type(c_ptr), value) - Required - Stop event handle.
- **flags** (integer(c_int), value) - Required - Launch flags.

### Return Value
- **integer(kind(hipsuccess))** - Returns a HIP success code indicating the result of the launch.
```

--------------------------------

### rocblas_fill_upper packed layout example

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__rocblas_1_1rocblas__sspr2__batched.html

Shows how the packed array AP is laid out when uplo is rocblas_fill_upper. The upper triangular portion of each symmetric matrix is stored column-by-column, with an example for n = 4.

```text
if uplo == rocblas_fill_upper:
    The upper triangular portion of each symmetric matrix A_i is supplied.
    The matrix is compacted so that AP contains the triangular portion
    column-by-column
    so that:
    AP(0) = A(0,0)
    AP(1) = A(0,1)
    AP(2) = A(1,1), etc.
        Ex: (rocblas_fill_upper; n = 4)
            1 2 4 7
            2 3 5 8   -----> [1, 2, 3, 4, 5, 6, 7, 8, 9, 0]
            4 5 6 9
            7 8 9 0
```

--------------------------------

### rocblas_get_performance_metric

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort__rocblas_8F90.html

Get performance metric.

```APIDOC
interface hipfort_rocblas::rocblas_get_performance_metric
```

--------------------------------

### hipfort::hipextlaunchkernel

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort_8F90.html

Launches kernel from the pointer address, with arguments and shared memory on stream.

```APIDOC
## hipfort::hipextlaunchkernel

### Description
Launches kernel from the pointer address, with arguments and shared memory on stream.

### Interface
`hipfort::hipextlaunchkernel`

### Parameters
No parameters are documented in the source.

### Returns
No return value is documented in the source.
```

--------------------------------

### hipfftmakeplanmany

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/interfacehipfort__hipfft_1_1hipfftmakeplanmany.html

Initializes a new batched rank-dimensional FFT plan with advanced data layout.

```APIDOC
## Function: hipfftmakeplanmany

### Description
Initializes a new batched rank-dimensional FFT plan with advanced data layout. This function modifies an existing plan handle to configure the FFT transform parameters.

### Signature
integer(kind(hipfft_success)) function hipfftmakeplanmany_(plan, rank, n, inembed, istride, idist, onembed, ostride, odist, mytype, batch, worksize)

### Parameters
- **plan** (integer, out) - Pointer to the FFT plan handle.
- **rank** (integer, in) - Dimension of transform (1, 2, or 3).
- **n** (integer, in) - Number of elements to transform in the x/y/z directions.
- **inembed** (integer, in) - Number of elements in the input data in the x/y/z directions.
- **istride** (integer, in) - Distance between two successive elements in the input data.
- **idist** (integer, in) - Distance between input batches.
- **onembed** (integer, in) - Number of elements in the output data in the x/y/z directions.
- **ostride** (integer, in) - Distance between two successive elements in the output data.
- **odist** (integer, in) - Distance between output batches.
- **mytype** (integer, in) - FFT type.
- **batch** (integer, in) - Number of batched transforms to perform.
- **worksize** (integer, out) - Pointer to work area size (returned value).
```

--------------------------------

### rocblas_get_pointer_mode

Source: https://rocm.docs.amd.com/projects/hipfort/en/latest/doxygen/html/hipfort__rocblas_8F90.html

Get pointer mode.

```APIDOC
interface hipfort_rocblas::rocblas_get_pointer_mode
```
Tokens:
Raw