# HSA.jl

**Repository Path**: Julialang/HSA.jl

## Basic Information

- **Project Name**: HSA.jl
- **Description**: No description available
- **Primary Language**: Unknown
- **License**: MIT
- **Default Branch**: master
- **Homepage**: None
- **GVP Project**: No

## Statistics

- **Stars**: 0
- **Forks**: 0
- **Created**: 2018-03-12
- **Last Updated**: 2024-08-18

## Categories & Tags

**Categories**: Uncategorized

**Tags**: None

## README

# HSA.jl

This is a package wrapping the Heterogeneous System Architecture (HSA) runtime libraries.

For links to official HSA Documentation, check out the
[wiki](https://github.com/JuliaGPU/HSA.jl/wiki/HSA.jl)

### Library Discovery
The package needs to know where to find the HSA runtime.
The code responsible for discovering the HSA libraries and headers is located in
src/discovery.jl. It expects to find the path containing the HSA libraries
(libhsa-runtime64, libhsakmt [and libhsa-runtime-ext64.so]) in
`ENV["LD_LIBRARY_PATH"]`.

For a port of the `vector_copy` example that comes with the HSA runtime, see the
`sample` directory.

### Julia functions as Kernels
In combination with a [modified Julia
binary](https://github.com/rollingthunder/julia), HSA.jl can be used to execute
Kernel functions written in Julia on an HSA agent. 

Minimal example:

```julia
using HSA

@hsa_kernel function vector_copy(a, b)
    i = get_global_id()
    b[i] = a[i]
end

N =  1024*1024

a_in = Array(Int, N)
rand!(a_in)

b_out = Array(Int, N)
rand!(b_out)

assert(a_in != b_out)

@hsa (N) vector_copy_kernel(a_in, b_out)

assert(a_in == b_out)
```

More examples using this facility can be found in `samples/codegen`

### Module overview
*   HSA
    Container Module for all HSA functionality.
    +   Wrappers for functions and constants defined by the HSA runtime.
    +   Functionality to emulate execution of Julia kernels
        -   On a non-HSA CPU
        -   if there is no codegen support
*   HSA.Builtins
    Functions that compile down to device intrinsics for use in Kernels
*   HSA.ExtFinalization
    Custom Wrappers for the finalization extension to the HSA Runtime

### Source Folder
*   binding
    Contains the code that interfaces with the various parts of the HSA Runtime
    +   generated
        Files generated from the hsa headers via gen/generate.jl 
    +   custom
        Custom wrapper code 
*   codegen
    Contains all code that relies on the modified julia code generator to work
*   emulation
    Provides emulation for some parts of the codegen infrastructure even when no
    hsa or codegen are available

### Source Generation
The files in src/binding/generated are generated by the script gen/generate.jl.
Regenerating these is usually unnecessary unless the HSA headers have changed or
you made changes to the generation script.
The script has some additional dependencies (mainly Clang.jl) and needs to be able to
find the HSA headers.

### Known Issues
*   HSA.executable\_get\_symbol does not find symbols  
    Maybe the name string for the symbol is not being passed correctly?  
    Workaround: iterate over all symbols and inspect their names