Area HIP Implementation API

group aml_area_hip

Hip Implementation of Areas. 

#include <aml/area/hip.h>

Hip implementation of AML areas. This building block relies on Hip implementation of malloc/free to provide mmap/munmap on device memory.

AML hip areas may be created to allocate current or specific hip devices. Also allocations can be private to a single device or shared across devices. Finally allocations can be backed by host memory allocation.

Defines

AML_AREA_HIP_FLAG_DEFAULT

Default hip area flags.

  • Allocation on device only,

  • Allocation visible by a single device.

  • Allocation not mapped on host memory.

AML_AREA_HIP_FLAG_ALLOC_HOST

Device allocation flag. Default behaviour is allocation on device. If this flag is set then allocation will be on host.

AML_AREA_HIP_FLAG_ALLOC_UNIFIED

Unified memory flag. If this flag is set, then allocation will create a unified memory pointer usable on host and device. Additionally, AML_AREA_HIP_FLAG_ALLOC_HOST and AML_AREA_HIP_FLAG_ALLOC_MAPPED will be ignored.

See also

hipMallocManaged()

AML_AREA_HIP_FLAG_ALLOC_MAPPED

Mapping flag. Default behaviour is allocation not mapped. If set, the pointer returned by mmap function will be host side memory mapped on device. A pointer to device memory can then be retrieved by calling hipHostGetDevicePointer(). If AML_AREA_HIP_FLAG_ALLOC_HOST is set, then host side memory will be allocated. Else, “ptr” field of mmap options will be used to map device memory (“ptr” must not be NULL).

See also

hipHostRegister(), hipHostAlloc().

AML_AREA_HIP_FLAG_ALLOC_GLOBAL

Unified memory setting flag. If AML_AREA_HIP_FLAG_ALLOC_UNIFIED is set, then this flag is looked to set hipMallocManaged() flag hipAttachGlobal. Else if AML_AREA_HIP_FLAG_ALLOC_MAPPED is set, or AML_AREA_HIP_FLAG_ALLOC_HOST flag is set, then this flag is looked to set hipMallocHost() flag hipHostAllocPortable. The default behaviour is to make allocation visible from a single device. If this flag is set, then allocation will be visible on all devices.

See also

hipMallocManaged()

Functions

int aml_area_hip_create(struct aml_area **area, const int device, const int flags)

Hip area creation.

See also

AML_AREA_HIP_FLAG_*.

Parameters:

  • area – [out] pointer to an uninitialized struct aml_area pointer to receive the new area.

  • device – [in] A valid hip device id, i.e from 0 to num_devices-1. If device id is negative, then current hip device will be used using aml_area_hip_mmap().

  • flags – [in] Allocation flags.

Returns:

AML_SUCCESS on success and area points to the new aml_area.

Returns:

-AML_FAILURE if hip API failed to provide the number of devices.

Returns:

-AML_EINVAL if device id is greater than or equal to the number of devices.

Returns:

-AML_ENOMEM if space to carry area cannot be allocated.

void aml_area_hip_destroy(struct aml_area **area)

Hip area destruction.

Destroy (finalize and free resources) a struct aml_area created by aml_area_hip_create().

Parameters:

area – [inout] is NULL after this call.

void *aml_area_hip_mmap(const struct aml_area_data *area_data, size_t size, struct aml_area_mmap_options *options)

Hip implementation of mmap operation for aml area.

This function is a wrapper on hip alloc functions. It uses area settings to: select device on which to perform allocation, select allocation function and set its parameters. Any pointer obtained through aml_area_hip_mmap() must be unmapped with aml_area_hip_munmap().

Device selection is not thread safe and requires to set the global state of hip library. When selecting a device, allocation may succeed while setting device back to original context devices may fail. In that case, you need to set aml_errno to AML_SUCCESS prior to calling this function in order to catch the error when return value is not NULL.

See also

AML_AREA_HIP_FLAG_*

Parameters:

  • area_data – [in] The structure containing hip area settings.

  • size – [in] The size to allocate.

  • options – [in] A struct aml_area_hip_mmap_options *. If > 0, device will be used to select the target device. If area flags AML_AREA_HIP_FLAG_MAPPED is set and AML_AREA_HIP_FLAG_HOST is not set, then options field “ptr” must not be NULL and point to a host memory that can be mapped on GPU.

Returns:

NULL on failure with aml errno set to the following error codes: AML_ENOTSUP is one of the hip calls failed with error: hipErrorInsufficientDriver, hipErrorNoDevice.

  • AML_EINVAL if target device id is not valid or provided argument are not compatible.

  • AML_EBUSY if a specific device was requested but was in already use.

  • AML_ENOMEM if memory allocation failed with error hipErrorMemoryAllocation.

  • AML_FAILURE if one of the hip calls resulted in error hipErrorInitializationError.

Returns:

A hip pointer usable on device and host if area flags contains AML_AREA_HIP_FLAG_ALLOC_UNIFIED.

Returns:

A pointer to host memory on which one can call hipHostGetDevicePointer() to get a pointer to mapped device memory, if AML_AREA_HIP_FLAG_ALLOC_MAPPED is set. Obtained pointer must be unmapped with aml_area_hip_munmap(). If host side memory was provided as mmap option, then it still has to be freed.

Returns:

A pointer to host memory if area flag AML_AREA_HIP_FLAG_ALLOC_HOST is set.

Returns:

A pointer to device memory if no flag is set.

int aml_area_hip_munmap(const struct aml_area_data *area_data, void *ptr, const size_t size)

munmap hook for aml area.

unmap memory mapped with aml_area_hip_mmap().

Parameters:

  • area_data – [in] Ignored

  • ptr – [inout] The virtual memory to unmap.

  • size – [in] The size of virtual memory to unmap.

Returns:

-AML_EINVAL if hipFree() returned hipErrorInvalidValue.

Returns:

AML_SUCCESS otherwise.

Variables

struct aml_area_ops aml_area_hip_ops

aml area hooks for hip implementation.

struct aml_area aml_area_hip

Default hip area: Allocation on device, visible by a single device, and not mapped on host memory.

struct aml_area aml_area_hip_unified

Hip area allocating unified memory.

struct aml_area_hip_mmap_options
#include <hip.h>

Options that can eventually be passed to mmap call.

struct aml_area_hip_data
#include <hip.h>

Implementation of aml_area_data for hip areas.