device_management_hip.cpp File Reference

#include "gmxpre.h"
#include "config.h"
#include <algorithm>
#include <optional>
#include <vector>
#include <hip/hip_runtime_api.h>
#include "gromacs/gpu_utils/device_context.h"
#include "gromacs/gpu_utils/device_stream.h"
#include "gromacs/gpu_utils/hiputils.h"
#include "gromacs/utility/exceptions.h"
#include "gromacs/utility/logger.h"
#include "gromacs/utility/mpiinfo.h"
#include "gromacs/utility/programcontext.h"
#include "gromacs/utility/stringutil.h"
#include "device_information.h"
#include "device_management.h"
#include "device_management_shared_amd.h"

Include dependency graph for device_management_hip.cpp:

Description

Defines the HIP implementations of the device management.

Author

Anca Hamuraru anca@.nosp@m.stre.nosp@m.amcom.nosp@m.puti.nosp@m.ng.eu

Dimitrios Karkoulis dimit.nosp@m.ris..nosp@m.karko.nosp@m.ulis.nosp@m.@gmai.nosp@m.l.co.nosp@m.m

Teemu Virolainen teemu.nosp@m.@str.nosp@m.eamco.nosp@m.mput.nosp@m.ing.e.nosp@m.u

Mark Abraham mark..nosp@m.j.ab.nosp@m.raham.nosp@m.@gma.nosp@m.il.co.nosp@m.m

Szilárd Páll pall..nosp@m.szil.nosp@m.ard@g.nosp@m.mail.nosp@m..com

Artem Zhmurov zhmur.nosp@m.ov@g.nosp@m.mail..nosp@m.com

Paul Bauer paul..nosp@m.baue.nosp@m.r.q@g.nosp@m.mail.nosp@m..com

Functions
static __global__ void	dummy_kernel ()
	Dummy kernel used for sanity checking. More...
void	warnWhenDeviceNotTargeted (const gmx::MDLogger &mdlog, const DeviceInformation &deviceInfo)
	Warn to the logger when the detected device was not one of the targets selected at configure time for compilation. More...
static DeviceStatus	checkDevicePairlistCompatible (const DeviceInformation deviceInfo)
	Checks that device deviceInfo is compatible with GROMACS pairlist layout.. More...
static bool	determineIfDeviceHasLargeRegisterPool (std::string deviceArch)
static DeviceStatus	checkDeviceStatus (const DeviceInformation &deviceInfo)
	Runs GPU compatibility and sanity checks on the indicated device. More...
bool	isDeviceDetectionFunctional (std::string *errorMessage)
	Return whether GPU detection is functioning correctly. More...
std::vector< std::unique_ptr < DeviceInformation > >	findDevices ()
	Find all GPUs in the system. More...
void	setActiveDevice (const DeviceInformation &deviceInfo)
	Set the active GPU. More...
void	releaseDevice ()
	Releases the GPU device used by the active context at the time of calling. More...
std::string	getDeviceInformationString (const DeviceInformation &deviceInfo)
	Formats and returns a device information string for a given GPU. More...

Variables
static const std::vector < std::string >	archsWithLargeRegisterPool
	List of known architectures with large register pool. More...

Function Documentation

static DeviceStatus checkDevicePairlistCompatible ( const DeviceInformation  deviceInfo )

static

Checks that device deviceInfo is compatible with GROMACS pairlist layout..

Parameters

[in]

deviceInfo

HIP device information.

Returns

The status enumeration value for the checked device.

static DeviceStatus checkDeviceStatus ( const DeviceInformation &  deviceInfo )

static

Runs GPU compatibility and sanity checks on the indicated device.

Runs a series of checks to determine that the given GPU and underlying HIP driver/runtime functions properly.

As the error handling only permits returning the state of the GPU, this function does not clear the HIP runtime API status allowing the caller to inspect the error upon return. Note that this also means it is the caller's responsibility to reset the HIP runtime state.

Todo:

Currently we do not make a distinction between the type of errors that can appear during functionality checks. This needs to be improved, e.g if the dummy test kernel fails to execute with a "device busy message" we should appropriately report that the device is busy instead of NonFunctional.

Todo:

Introduce errors codes and handle errors more smoothly.

Parameters

[in]

deviceInfo

Device information on the device to check.

Returns

The status enumeration value for the checked device:

static __global__ void dummy_kernel ( )

static

Dummy kernel used for sanity checking.

std::vector<std::unique_ptr<DeviceInformation> > findDevices

(

)

Find all GPUs in the system.

Will detect every GPU supported by the device driver in use. Must only be called if canPerformDeviceDetection() has returned true. This routine also checks for the compatibility of each device and fill the deviceInfo array with the required information on each device: ID, device properties, status.

Note that this function leaves the GPU runtime API error state clean; this is implemented ATM in the CUDA flavor. This invalidates any existing CUDA streams, allocated memory on GPU, etc.

Todo:

: Check if errors do propagate in OpenCL as they do in CUDA and whether there is a mechanism to "clear" them.

Returns

Standard vector with the list of devices found

Exceptions

InternalError

if a GPU API returns an unexpected failure (because the call to canDetectGpus() should always prevent this occuring)

std::string getDeviceInformationString

(

const DeviceInformation &

deviceInfo

)

Formats and returns a device information string for a given GPU.

Given an index directly into the array of available GPUs, returns a formatted info string for the respective GPU which includes ID, name, compute capability, and detection status.

Parameters

[in]

deviceInfo

An information on device that is to be set.

Returns

A string describing the device.

bool isDeviceDetectionFunctional

(

std::string *

errorMessage

)

Return whether GPU detection is functioning correctly.

Returns true when this is a build of GROMACS configured to support GPU usage, and a valid device driver, ICD, and/or runtime was detected.

This function is not intended to be called from build configurations that do not support GPUs, and there will be no descriptive message in that case.

Parameters

[out]

errorMessage

When returning false on a build configured with GPU support and non-nullptr was passed, the string contains a descriptive message about why GPUs cannot be detected.

Does not throw.

void releaseDevice

(

)

Releases the GPU device used by the active context at the time of calling.

With CUDA, the device is reset and therefore all data uploaded to the GPU is lost. This must only be called when none of this data is required anymore, because subsequent attempts to free memory associated with the context will otherwise fail. Calls gmx_warning upon errors.

With other GPU SDKs, does nothing.

Should only be called after setActiveDevice was called.

void setActiveDevice

(

const DeviceInformation &

deviceInfo

)

Set the active GPU.

This sets the device for which the device information is passed active. Essential in CUDA, where the device buffers and kernel launches are not connected to the device context. In OpenCL, checks the device vendor and makes vendor-specific performance adjustments.

Parameters

[in]

deviceInfo

Information on the device to be set.

Issues a fatal error for any critical errors that occur during initialization.

void warnWhenDeviceNotTargeted	(	const gmx::MDLogger &	mdlog,
		const DeviceInformation &	deviceInfo
	)

Warn to the logger when the detected device was not one of the targets selected at configure time for compilation.

Parameters

[in]	mdlog	Logger
[in]	deviceInfo	The device to potentially warn about

Variable Documentation

const std::vector<std::string> archsWithLargeRegisterPool

static

Initial value:

= { "gfx90a",
                                                                     "gfx940",
                                                                     "gfx941",
                                                                     "gfx942" }

List of known architectures with large register pool.