IGPUKernelSizeOptimizerSvc Class Reference

Actual Athena inteface for the IGPUKernelSizeOptimizer. More...

#include <IGPUKernelSizeOptimizerSvc.h>

Inheritance diagram for IGPUKernelSizeOptimizerSvc:

Collaboration diagram for IGPUKernelSizeOptimizerSvc:

Public Types
enum	SpecialSizeHints { CooperativeLaunch = -1 }

Public Member Functions
	DeclareInterfaceID (IGPUKernelSizeOptimizerSvc, 1, 0)
virtual	~IGPUKernelSizeOptimizerSvc ()=default
virtual void	register_kernel (const std::string &kernel_name, void *kernel, const int blocksize_hint, const int gridsize_hint, const int max_total_threads)
	Register a kernel with a specific name. More...
virtual void	register_kernels (const std::string &tool_name, const int number, void **kernels, const int *blocksize_hints, const int *gridsize_hints, const int *max_total_threads, const int offset=0)=0
	Register a set of kernels that can be referred back to with a name and a number. More...
virtual CUDAKernelLaunchConfiguration	get_launch_configuration (const std::string &name, const int number=0, const int dynamic_memory=0) const =0
	Retrieve the (hopefully optimal) kernel launch configuration. More...
virtual bool	can_use_cooperative_groups () const =0
	Whether the device + environment in use support cooperative groups. More...
virtual bool	can_use_dynamic_parallelism () const =0
	Whether the device + environment in use support dynamic parallelism. More...
virtual bool	use_minimal_kernel_sizes () const
	Whether to avoid oversizing kernels and instead (if possible) launch kernels with the exact number of threads... More...

Detailed Description

Actual Athena inteface for the IGPUKernelSizeOptimizer.

Author

Nuno Fernandes nuno..nosp@m.dos..nosp@m.santo.nosp@m.s.fe.nosp@m.rnand.nosp@m.es@c.nosp@m.ern.c.nosp@m.h

Date

04 August 2023

Definition at line 20 of file IGPUKernelSizeOptimizerSvc.h.

Member Enumeration Documentation

SpecialSizeHints

enum IGPUKernelSizeOptimizer::SpecialSizeHints

inherited

Enumerator
CooperativeLaunch

Definition at line 32 of file IGPUKernelSizeOptimizer.h.

  {
    CooperativeLaunch = -1
  };

Constructor & Destructor Documentation

~IGPUKernelSizeOptimizerSvc()

virtual IGPUKernelSizeOptimizerSvc::~IGPUKernelSizeOptimizerSvc ( )

virtualdefault

Member Function Documentation

can_use_cooperative_groups()

virtual bool IGPUKernelSizeOptimizer::can_use_cooperative_groups ( ) const

pure virtualinherited

Whether the device + environment in use support cooperative groups.

can_use_dynamic_parallelism()

virtual bool IGPUKernelSizeOptimizer::can_use_dynamic_parallelism ( ) const

pure virtualinherited

Whether the device + environment in use support dynamic parallelism.

DeclareInterfaceID()

IGPUKernelSizeOptimizerSvc::DeclareInterfaceID	(	IGPUKernelSizeOptimizerSvc	,
		1	,
		0
	)

get_launch_configuration()

virtual CUDAKernelLaunchConfiguration IGPUKernelSizeOptimizer::get_launch_configuration ( const std::string &  name, const int  number = 0, const int  dynamic_memory = 0  ) const

pure virtualinherited

Retrieve the (hopefully optimal) kernel launch configuration.

register_kernel()

virtual void IGPUKernelSizeOptimizer::register_kernel ( const std::string &  kernel_name, void *  kernel, const int  blocksize_hint, const int  gridsize_hint, const int  max_total_threads  )

inlinevirtualinherited

Register a kernel with a specific name.

Definition at line 39 of file IGPUKernelSizeOptimizer.h.

  {
    this->register_kernels(kernel_name, 1, &kernel, &blocksize_hint, &gridsize_hint, &max_total_threads, 0);
  }

register_kernels()

virtual void IGPUKernelSizeOptimizer::register_kernels ( const std::string &  tool_name, const int  number, void **  kernels, const int *  blocksize_hints, const int *  gridsize_hints, const int *  max_total_threads, const int  offset = 0  )

pure virtualinherited

Register a set of kernels that can be referred back to with a name and a number.

Uses C-style arrays for more immediate CUDA compatibility, assumes the size of kernels, blocksize_hints and gridsize_hints is number, and starts the numbering with an optional offset.

use_minimal_kernel_sizes()

virtual bool IGPUKernelSizeOptimizer::use_minimal_kernel_sizes ( ) const

inlinevirtualinherited

Whether to avoid oversizing kernels and instead (if possible) launch kernels with the exact number of threads...

Definition at line 74 of file IGPUKernelSizeOptimizer.h.

  {
    //Testing shows that, at least on the devices we use,
    //we only lose performance by dyn-par'ing our way to do this.
    return false;
  }

---

The documentation for this class was generated from the following file:

• IGPUKernelSizeOptimizerSvc.h