clsim C++ API Reference

class Axes : private noncopyable

#include <Axes.h>

The Axes encapsulates the coordinate system used for binning photon paths.

Subclassed by clsim::tabulator::CylindricalAxes, clsim::tabulator::SphericalAxes

Public Types

typedef boost::shared_ptr<Axis> value_type

Public Functions

Axes(const std::vector<value_type>&)

virtual ~Axes()

inline const value_type &at(unsigned i) const

inline size_t GetNDim() const

inline size_t GetNBins() const

Total number of bins in the bin content array.

inline std::vector<size_t> GetShape() const

Number of bins in each dimension.

inline std::vector<size_t> GetStrides() const

Offset between adjacent bins in each dimension.

std::string GenerateBinningCode() const

Generate the OpenCL code required to transform photon positions into source-relative positions and calculate the corresponding index in the flattened bin-content array.

virtual double GetBinVolume(const std::vector<size_t> &multiIndex) const = 0

Calculate the volume (in m^3) of the cell at index multiIndex in the bin-content array.

Protected Functions

virtual std::string GetCoordinateFunction() const = 0

Generate an OpenCL function that calculates the source-relative coordinates from the photon position and time.

virtual std::string GetBoundsCheckFunction() const = 0

Generate an OpenCL function that returns true if the photon has exited the recording volume and should be stopped.

Private Functions

std::string GetBinIndexFunction() const

Generate an OpenCL function that transforms a set of coordinates into an index in the flattened bin content array. This is synthesized from the properties of the individual Axis instances.

Private Members

std::vector<value_type> axes_

size_t n_dim_

size_t n_bins_

std::vector<size_t> shape_

std::vector<size_t> strides_

class Axis

#include <Axis.h>

Subclassed by clsim::tabulator::LinearAxis, clsim::tabulator::PowerAxis

Public Functions

Axis(double min, double max, unsigned n_bins)

virtual ~Axis()

inline double GetMin() const

lowest bin edge

inline double GetMax() const

highest bin edge

inline unsigned GetNBins() const

the number of bins without under/overflow

std::string GetIndexCode(const std::string &varName) const

std::vector<double> GetBinEdges() const

double GetBinEdge(unsigned i) const

virtual double Transform(double value) const = 0

Transform value from linear into nonlinear space.

virtual double InverseTransform(double value) const = 0

Transform value from nonlinear into linear space.

virtual std::string GetTransformCode(const std::string &varName) const = 0

Generate OpenCL code to transform variable from linear into nonlinear space.

virtual std::string GetInverseTransformCode(const std::string &varName) const = 0

Generate OpenCL code to transform variable from nonlinear into linear space.

Private Members

double min_

double max_

unsigned n_bins_

struct BarrierData_t

class Buffer : public cl::Memory

#include <opencl.hpp>

Class interface for Buffer Memory Objects.

See Memory for details about copy semantics, etc.

See also

Memory

Subclassed by cl::BufferGL, cl::BufferRenderGL

Public Functions

inline Buffer(const Context &context, cl_mem_flags flags, size_type size, void *host_ptr = NULL, cl_int *err = NULL)

Constructs a Buffer in a specified context.

Wraps clCreateBuffer().

Parameters:

host_ptr – Storage to be used if the CL_MEM_USE_HOST_PTR flag was specified. Note alignment & exclusivity requirements.

inline Buffer(cl_mem_flags flags, size_type size, void *host_ptr = NULL, cl_int *err = NULL)

Constructs a Buffer in the default context.

Wraps clCreateBuffer().

See also

Context::getDefault()

Parameters:

host_ptr – Storage to be used if the CL_MEM_USE_HOST_PTR flag was specified. Note alignment & exclusivity requirements.

template<typename IteratorType>
inline Buffer(IteratorType startIterator, IteratorType endIterator, bool readOnly, bool useHostPtr = false, cl_int *err = NULL)

Construct a Buffer from a host container via iterators. IteratorType must be random access. If useHostPtr is specified iterators must represent contiguous data.

template<typename IteratorType>
Buffer(const Context &context, IteratorType startIterator, IteratorType endIterator, bool readOnly, bool useHostPtr = false, cl_int *err = NULL)

Construct a Buffer from a host container via iterators using a specified context. IteratorType must be random access. If useHostPtr is specified iterators must represent contiguous data.

template<typename IteratorType>
Buffer(const CommandQueue &queue, IteratorType startIterator, IteratorType endIterator, bool readOnly, bool useHostPtr = false, cl_int *err = NULL)

Construct a Buffer from a host container via iterators using a specified queue. If useHostPtr is specified iterators must be random access.

inline Buffer()

Default constructor - initializes to NULL.

inline explicit Buffer(const cl_mem &buffer, bool retainObject = false)

Constructor from cl_mem - takes ownership.

See Memory for further details.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline Buffer &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Buffer(const Buffer &buf)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Buffer &operator=(const Buffer &buf)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Buffer (Buffer &&buf) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Buffer &operator=(Buffer &&buf)

Move assignment to forward move to the superclass correctly. Required for MSVC.

inline Buffer createSubBuffer(cl_mem_flags flags, cl_buffer_create_type buffer_create_type, const void *buffer_create_info, cl_int *err = NULL)

Creates a new buffer object from this.

Wraps clCreateSubBuffer().

class BufferGL : public cl::Buffer

#include <opencl.hpp>

Class interface for GL Buffer Memory Objects.

This is provided to facilitate interoperability with OpenGL.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline BufferGL(const Context &context, cl_mem_flags flags, cl_GLuint bufobj, cl_int *err = NULL)

Constructs a BufferGL in a specified context, from a given GL buffer.

Wraps clCreateFromGLBuffer().

inline BufferGL()

Default constructor - initializes to NULL.

inline explicit BufferGL(const cl_mem &buffer, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline BufferGL &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline BufferGL(const BufferGL &buf)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline BufferGL &operator=(const BufferGL &buf)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  BufferGL (BufferGL &&buf) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline BufferGL &operator=(BufferGL &&buf)

Move assignment to forward move to the superclass correctly. Required for MSVC.

inline cl_int getObjectInfo(cl_gl_object_type *type, cl_GLuint *gl_object_name)

Wrapper for clGetGLObjectInfo().

class BufferRenderGL : public cl::Buffer

#include <opencl.hpp>

Class interface for GL Render Buffer Memory Objects.

This is provided to facilitate interoperability with OpenGL.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline BufferRenderGL(const Context &context, cl_mem_flags flags, cl_GLuint bufobj, cl_int *err = NULL)

Constructs a BufferRenderGL in a specified context, from a given GL Renderbuffer.

Wraps clCreateFromGLRenderbuffer().

inline BufferRenderGL()

Default constructor - initializes to NULL.

inline explicit BufferRenderGL(const cl_mem &buffer, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline BufferRenderGL &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline BufferRenderGL(const BufferRenderGL &buf)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline BufferRenderGL &operator=(const BufferRenderGL &buf)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  BufferRenderGL (BufferRenderGL &&buf) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline BufferRenderGL &operator=(BufferRenderGL &&buf)

Move assignment to forward move to the superclass correctly. Required for MSVC.

inline cl_int getObjectInfo(cl_gl_object_type *type, cl_GLuint *gl_object_name)

Wrapper for clGetGLObjectInfo().

struct CascadeStepData_t

Public Members

I3Particle particle

uint32_t particleIdentifier

uint64_t photonsPerStep

uint64_t numSteps

uint64_t numPhotonsInLastStep

double pa

double pb

class CommandQueue : public cl::detail::Wrapper<cl_command_queue>

#include <opencl.hpp>

CommandQueue interface for cl_command_queue.

Public Functions

inline CommandQueue(cl_command_queue_properties properties, cl_int *err = NULL)

Constructs a CommandQueue based on passed properties. Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline CommandQueue(QueueProperties properties, cl_int *err = NULL)

Constructs a CommandQueue based on passed properties. Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline explicit CommandQueue(const Context &context, cl_command_queue_properties properties = 0, cl_int *err = NULL)

Constructs a CommandQueue for an implementation defined device in the given context Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline explicit CommandQueue(const Context &context, QueueProperties properties, cl_int *err = NULL)

Constructs a CommandQueue for an implementation defined device in the given context Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline CommandQueue(const Context &context, const Device &device, cl_command_queue_properties properties = 0, cl_int *err = NULL)

Constructs a CommandQueue for a passed device and context Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline CommandQueue(const Context &context, const Device &device, QueueProperties properties, cl_int *err = NULL)

Constructs a CommandQueue for a passed device and context Will return an CL_INVALID_QUEUE_PROPERTIES error if CL_QUEUE_ON_DEVICE is specified.

inline CommandQueue()

inline explicit CommandQueue(const cl_command_queue &commandQueue, bool retainObject = false)

Constructor from cl_command_queue - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline CommandQueue &operator=(const cl_command_queue &rhs)

inline CommandQueue(const CommandQueue &queue)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline CommandQueue &operator=(const CommandQueue &queue)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  CommandQueue (CommandQueue &&queue) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline CommandQueue &operator=(CommandQueue &&queue)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_command_queue_info name, T *param) const

template<cl_command_queue_info name>
inline detail::param_traits<detail::cl_command_queue_info, name>::param_type getInfo(cl_int *err = NULL) const

inline cl_int enqueueReadBuffer(const Buffer &buffer, cl_bool blocking, size_type offset, size_type size, void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueWriteBuffer(const Buffer &buffer, cl_bool blocking, size_type offset, size_type size, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueCopyBuffer(const Buffer &src, const Buffer &dst, size_type src_offset, size_type dst_offset, size_type size, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueReadBufferRect(const Buffer &buffer, cl_bool blocking, const array<size_type, 3> &buffer_offset, const array<size_type, 3> &host_offset, const array<size_type, 3> &region, size_type buffer_row_pitch, size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueWriteBufferRect(const Buffer &buffer, cl_bool blocking, const array<size_type, 3> &buffer_offset, const array<size_type, 3> &host_offset, const array<size_type, 3> &region, size_type buffer_row_pitch, size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueCopyBufferRect(const Buffer &src, const Buffer &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, size_type src_row_pitch, size_type src_slice_pitch, size_type dst_row_pitch, size_type dst_slice_pitch, const vector<Event> *events = NULL, Event *event = NULL) const

template<typename PatternType>
inline cl_int enqueueFillBuffer(const Buffer &buffer, PatternType pattern, size_type offset, size_type size, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueue a command to fill a buffer object with a pattern of a given size. The pattern is specified as a vector type.

Template Parameters:

• PatternType – The datatype of the pattern field. The pattern type must be an accepted OpenCL data type.

• offset – Is the offset in bytes into the buffer at which to start filling. This must be a multiple of the pattern size.

• size – Is the size in bytes of the region to fill. This must be a multiple of the pattern size.

inline cl_int enqueueReadImage(const Image &image, cl_bool blocking, const array<size_type, 3> &origin, const array<size_type, 3> &region, size_type row_pitch, size_type slice_pitch, void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueWriteImage(const Image &image, cl_bool blocking, const array<size_type, 3> &origin, const array<size_type, 3> &region, size_type row_pitch, size_type slice_pitch, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueCopyImage(const Image &src, const Image &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueFillImage(const Image &image, cl_float4 fillColor, const array<size_type, 3> &origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueue a command to fill an image object with a specified color.

Parameters:

fillColor – is the color to use to fill the image. This is a four component RGBA floating-point color value if the image channel data type is not an unnormalized signed or unsigned data type.

inline cl_int enqueueFillImage(const Image &image, cl_int4 fillColor, const array<size_type, 3> &origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueue a command to fill an image object with a specified color.

Parameters:

fillColor – is the color to use to fill the image. This is a four component RGBA signed integer color value if the image channel data type is an unnormalized signed integer type.

inline cl_int enqueueFillImage(const Image &image, cl_uint4 fillColor, const array<size_type, 3> &origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueue a command to fill an image object with a specified color.

Parameters:

fillColor – is the color to use to fill the image. This is a four component RGBA unsigned integer color value if the image channel data type is an unnormalized unsigned integer type.

inline cl_int enqueueCopyImageToBuffer(const Image &src, const Buffer &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &region, size_type dst_offset, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueCopyBufferToImage(const Buffer &src, const Image &dst, size_type src_offset, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL) const

inline void *enqueueMapBuffer(const Buffer &buffer, cl_bool blocking, cl_map_flags flags, size_type offset, size_type size, const vector<Event> *events = NULL, Event *event = NULL, cl_int *err = NULL) const

inline void *enqueueMapImage(const Image &buffer, cl_bool blocking, cl_map_flags flags, const array<size_type, 3> &origin, const array<size_type, 3> &region, size_type *row_pitch, size_type *slice_pitch, const vector<Event> *events = NULL, Event *event = NULL, cl_int *err = NULL) const

template<typename T>
inline cl_int enqueueMapSVM(T *ptr, cl_bool blocking, cl_map_flags flags, size_type size, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a raw SVM pointer.

template<typename T, class D>
inline cl_int enqueueMapSVM(cl::pointer<T, D> &ptr, cl_bool blocking, cl_map_flags flags, size_type size, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a cl::pointer instance.

template<typename T, class Alloc>
inline cl_int enqueueMapSVM(cl::vector<T, Alloc> &container, cl_bool blocking, cl_map_flags flags, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a cl::vector instance.

inline cl_int enqueueUnmapMemObject(const Memory &memory, void *mapped_ptr, const vector<Event> *events = NULL, Event *event = NULL) const

template<typename T>
inline cl_int enqueueUnmapSVM(T *ptr, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a raw SVM pointer.

template<typename T, class D>
inline cl_int enqueueUnmapSVM(cl::pointer<T, D> &ptr, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a cl::pointer instance.

template<typename T, class Alloc>
inline cl_int enqueueUnmapSVM(cl::vector<T, Alloc> &container, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a cl::vector instance.

inline cl_int enqueueMarkerWithWaitList(const vector<Event> *events = 0, Event *event = 0) const

Enqueues a marker command which waits for either a list of events to complete, or all previously enqueued commands to complete.

Enqueues a marker command which waits for either a list of events to complete, or if the list is empty it waits for all commands previously enqueued in command_queue to complete before it completes. This command returns an event which can be waited on, i.e. this event can be waited on to insure that all events either in the event_wait_list or all previously enqueued commands, queued before this command to command_queue, have completed.

inline cl_int enqueueBarrierWithWaitList(const vector<Event> *events = 0, Event *event = 0) const

A synchronization point that enqueues a barrier operation.

Enqueues a barrier command which waits for either a list of events to complete, or if the list is empty it waits for all commands previously enqueued in command_queue to complete before it completes. This command blocks command execution, that is, any following commands enqueued after it do not execute until it completes. This command returns an event which can be waited on, i.e. this event can be waited on to insure that all events either in the event_wait_list or all previously enqueued commands, queued before this command to command_queue, have completed.

inline cl_int enqueueMigrateMemObjects(const vector<Memory> &memObjects, cl_mem_migration_flags flags, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command to indicate with which device a set of memory objects should be associated.

template<typename T>
inline cl_int enqueueMigrateSVM(const cl::vector<T*> &svmRawPointers, const cl::vector<size_type> &sizes, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate ranges within a set of SVM allocations with a device.

Parameters:

sizes – - The length from each pointer to migrate.

template<typename T>
inline cl_int enqueueMigrateSVM(const cl::vector<T*> &svmRawPointers, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate a set of SVM allocations with a device.

template<typename T, class D>
inline cl_int enqueueMigrateSVM(const cl::vector<cl::pointer<T, D>> &svmPointers, const cl::vector<size_type> &sizes, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate ranges within a set of SVM allocations with a device.

Parameters:

sizes – - The length from each pointer to migrate.

template<typename T, class D>
inline cl_int enqueueMigrateSVM(const cl::vector<cl::pointer<T, D>> &svmPointers, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate a set of SVM allocations with a device.

template<typename T, class Alloc>
inline cl_int enqueueMigrateSVM(const cl::vector<cl::vector<T, Alloc>> &svmContainers, const cl::vector<size_type> &sizes, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate ranges within a set of SVM allocations with a device.

Parameters:

sizes – - The length from the beginning of each container to migrate.

template<typename T, class Alloc>
inline cl_int enqueueMigrateSVM(const cl::vector<cl::vector<T, Alloc>> &svmContainers, cl_mem_migration_flags flags = 0, const vector<Event> *events = NULL, Event *event = NULL) const

Enqueues a command that will allow the host associate a set of SVM allocations with a device.

inline cl_int enqueueNDRangeKernel(const Kernel &kernel, const NDRange &offset, const NDRange &global, const NDRange &local = NullRange, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int enqueueNativeKernel (void(CL_CALLBACK *userFptr)(void *), std::pair< void *, size_type > args, const vector< Memory > *mem_objects=NULL, const vector< const void * > *mem_locs=NULL, const vector< Event > *events=NULL, Event *event=NULL) const

inline cl_int enqueueAcquireGLObjects(const vector<Memory> *mem_objects = NULL, const vector<Event> *events = NULL, Event *event = NULL) const

Deprecated APIs for 1.2

inline cl_int enqueueReleaseGLObjects(const vector<Memory> *mem_objects = NULL, const vector<Event> *events = NULL, Event *event = NULL) const

inline cl_int flush() const

Deprecated APIs for 1.2

inline cl_int finish() const

Public Static Functions

static inline CommandQueue getDefault(cl_int *err = NULL)

static inline CommandQueue setDefault(const CommandQueue &default_queue)

Modify the default command queue to be used by subsequent operations. Will only set the default if no default was previously created.

Returns:

updated default command queue. Should be compared to the passed value to ensure that it was updated.

Private Static Functions

static inline void makeDefault()

Create the default command queue returned by getDefault.

It sets default_error_ to indicate success or failure. It does not throw cl::Error.

static inline void makeDefaultProvided(const CommandQueue &c)

Create the default command queue.

This sets default_. It does not throw cl::Error.

Private Static Attributes

static std::once_flag default_initialized_

static CommandQueue default_

static cl_int default_error_ = CL_SUCCESS

class Context : public cl::detail::Wrapper<cl_context>

#include <opencl.hpp>

Class interface for cl_context.

Deprecated APIs for 1.2

See also

cl_context

Note

Copies of these objects are shallow, meaning that the copy will refer to the same underlying cl_context as the original. For details, see clRetainContext() and clReleaseContext().

Public Functions

inline  Context (const vector< Device > &devices, const cl_context_properties *properties=NULL, void(CL_CALLBACK *notifyFptr)(const char *, const void *, size_type, void *)=NULL, void *data=NULL, cl_int *err=NULL)

Constructs a context including a list of specified devices.

Wraps clCreateContext().

inline  Context (const Device &device, const cl_context_properties *properties=NULL, void(CL_CALLBACK *notifyFptr)(const char *, const void *, size_type, void *)=NULL, void *data=NULL, cl_int *err=NULL)

Constructs a context including a specific device.

Wraps clCreateContext().

inline  Context (cl_device_type type, const cl_context_properties *properties=NULL, void(CL_CALLBACK *notifyFptr)(const char *, const void *, size_type, void *)=NULL, void *data=NULL, cl_int *err=NULL)

Constructs a context including all or a subset of devices of a specified type.

Wraps clCreateContextFromType().

inline Context(const Context &ctx)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Context &operator=(const Context &ctx)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Context (Context &&ctx) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Context &operator=(Context &&ctx)

Move assignment to forward move to the superclass correctly. Required for MSVC.

inline Context()

Default constructor - initializes to NULL.

inline explicit Context(const cl_context &context, bool retainObject = false)

Constructor from cl_context - takes ownership.

This effectively transfers ownership of a refcount on the cl_context into the new Context object.

inline Context &operator=(const cl_context &rhs)

Assignment operator from cl_context - takes ownership.

This effectively transfers ownership of a refcount on the rhs and calls clReleaseContext() on the value previously held by this instance.

template<typename T>
inline cl_int getInfo(cl_context_info name, T *param) const

Wrapper for clGetContextInfo().

template<cl_context_info name>
inline detail::param_traits<detail::cl_context_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetContextInfo() that returns by value.

inline cl_int getSupportedImageFormats(cl_mem_flags flags, cl_mem_object_type type, vector<ImageFormat> *formats) const

Gets a list of supported image formats.

Wraps clGetSupportedImageFormats().

Public Static Functions

static inline Context getDefault(cl_int *err = NULL)

Returns a singleton context including all devices of CL_DEVICE_TYPE_DEFAULT.

Note

All calls to this function return the same cl_context as the first.

static inline Context setDefault(const Context &default_context)

Modify the default context to be used by subsequent operations. Will only set the default if no default was previously created.

Returns:

updated default context. Should be compared to the passed value to ensure that it was updated.

Private Static Functions

static inline void makeDefault()

Create the default context from the default device type in the default platform.

This sets default_ and default_error_. It does not throw cl::Error.

static inline void makeDefaultProvided(const Context &c)

Create the default context from a provided Context.

This sets default_. It does not throw cl::Error.

Private Static Attributes

static std::once_flag default_initialized_

static Context default_

static cl_int default_error_ = CL_SUCCESS

class CylindricalAxes : public clsim::tabulator::Axes

#include <Axes.h>

[Half]-Cylindrical coordinate system centered on the source axis, appropriate for a source moving at the speed of light with infinite range. NB: since the position of the source is degenerate with time, the reciever depth is used as a coordinate instead of the source depth.

Public Functions

inline CylindricalAxes(const std::vector<value_type> &axes)

virtual double GetBinVolume(const std::vector<size_t> &multiIndex) const

Calculate the volume (in m^3) of the cell at index multiIndex in the bin-content array.

Protected Functions

virtual std::string GetCoordinateFunction() const

Generate an OpenCL function that calculates the source-relative coordinates from the photon position and time.

virtual std::string GetBoundsCheckFunction() const

Generate an OpenCL function that returns true if the photon has exited the recording volume and should be stopped.

template<class Alloc>
class Deleter

#include <opencl.hpp>

Public Types

typedef std::allocator_traits<Alloc>::pointer pointer

Public Functions

inline Deleter(const Alloc &alloc, size_type copies)

inline void operator()(pointer ptr) const

Private Members

Alloc alloc_

size_type copies_

class Device : public cl::detail::Wrapper<cl_device_id>

#include <opencl.hpp>

Class interface for cl_device_id.

See also

cl_device_id

Note

Copies of these objects are inexpensive, since they don’t ‘own’ any underlying resources or data structures.

Public Functions

inline Device()

Default constructor - initializes to NULL.

inline explicit Device(const cl_device_id &device, bool retainObject = false)

Constructor from cl_device_id.

This simply copies the device ID value, which is an inexpensive operation.

inline Device &operator=(const cl_device_id &rhs)

Assignment operator from cl_device_id.

This simply copies the device ID value, which is an inexpensive operation.

inline Device(const Device &dev)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Device &operator=(const Device &dev)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Device (Device &&dev) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Device &operator=(Device &&dev)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_device_info name, T *param) const

Wrapper for clGetDeviceInfo().

template<cl_device_info name>
inline detail::param_traits<detail::cl_device_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetDeviceInfo() that returns by value.

inline cl_ulong getHostTimer(cl_int *error = nullptr)

Return the current value of the host clock as seen by the device. The resolution of the device timer may be queried with the CL_DEVICE_PROFILING_TIMER_RESOLUTION query.

Returns:

The host timer value.

inline std::pair<cl_ulong, cl_ulong> getDeviceAndHostTimer(cl_int *error = nullptr)

Return a synchronized pair of host and device timestamps as seen by device. Use to correlate the clocks and get the host timer only using getHostTimer as a lower cost mechanism in between calls. The resolution of the host timer may be queried with the CL_PLATFORM_HOST_TIMER_RESOLUTION query. The resolution of the device timer may be queried with the CL_DEVICE_PROFILING_TIMER_RESOLUTION query.

Returns:

A pair of (device timer, host timer) timer values.

inline cl_int createSubDevices(const cl_device_partition_property *properties, vector<Device> *devices)

Wrapper for clCreateSubDevices().

CL 1.2 version

Public Static Functions

static inline Device getDefault(cl_int *errResult = NULL)

Returns the first device on the default context.

See also

Context::getDefault()

static inline Device setDefault(const Device &default_device)

Modify the default device to be used by subsequent operations. Will only set the default if no default was previously created.

Returns:

updated default device. Should be compared to the passed value to ensure that it was updated.

Private Static Functions

static inline void makeDefault()

Create the default context.

This sets default_ and default_error_. It does not throw cl::Error.

static inline void makeDefaultProvided(const Device &p)

Create the default platform from a provided platform.

This sets default_. It does not throw cl::Error.

Private Static Attributes

static std::once_flag default_initialized_

static Device default_

static cl_int default_error_ = CL_SUCCESS

class DeviceCommandQueue : public cl::detail::Wrapper<cl_command_queue>

#include <opencl.hpp>

DeviceCommandQueue interface for device cl_command_queues.

Public Functions

inline DeviceCommandQueue()

Trivial empty constructor to create a null queue.

inline DeviceCommandQueue(DeviceQueueProperties properties, cl_int *err = NULL)

Default construct device command queue on default context and device

inline DeviceCommandQueue(const Context &context, const Device &device, DeviceQueueProperties properties = DeviceQueueProperties::None, cl_int *err = NULL)

Create a device command queue for a specified device in the passed context.

inline DeviceCommandQueue(const Context &context, const Device &device, cl_uint queueSize, DeviceQueueProperties properties = DeviceQueueProperties::None, cl_int *err = NULL)

Create a device command queue for a specified device in the passed context.

inline explicit DeviceCommandQueue(const cl_command_queue &commandQueue, bool retainObject = false)

Constructor from cl_command_queue - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline DeviceCommandQueue &operator=(const cl_command_queue &rhs)

inline DeviceCommandQueue(const DeviceCommandQueue &queue)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline DeviceCommandQueue &operator=(const DeviceCommandQueue &queue)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  DeviceCommandQueue (DeviceCommandQueue &&queue) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline DeviceCommandQueue &operator=(DeviceCommandQueue &&queue)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_command_queue_info name, T *param) const

template<cl_command_queue_info name>
inline detail::param_traits<detail::cl_command_queue_info, name>::param_type getInfo(cl_int *err = NULL) const

Public Static Functions

static inline DeviceCommandQueue makeDefault(cl_int *err = nullptr)

Create a new default device command queue for the default device, in the default context and of the default size. If there is already a default queue for the specified device this function will return the pre-existing queue.

static inline DeviceCommandQueue makeDefault(const Context &context, const Device &device, cl_int *err = nullptr)

Create a new default device command queue for the specified device and of the default size. If there is already a default queue for the specified device this function will return the pre-existing queue.

static inline DeviceCommandQueue makeDefault(const Context &context, const Device &device, cl_uint queueSize, cl_int *err = nullptr)

Create a new default device command queue for the specified device and of the requested size in bytes. If there is already a default queue for the specified device this function will return the pre-existing queue.

static inline DeviceCommandQueue updateDefault(const Context &context, const Device &device, const DeviceCommandQueue &default_queue, cl_int *err = nullptr)

Modify the default device command queue to be used for subsequent kernels. This can update the default command queue for a device repeatedly to account for kernels that rely on the default.

Returns:

updated default device command queue.

static inline DeviceCommandQueue getDefault(const CommandQueue &queue, cl_int *err = NULL)

Return the current default command queue for the specified command queue

struct domStruct

Public Members

unsigned int domID

double posX

double posY

double posZ

double cableAngle

template<typename CBRNG>
struct Engine

#include <Engine.hpp>

If G satisfies the requirements of a CBRNG, and has a ctr_type whose value_type is an unsigned integral type, then Engine<G> satisfies the requirements of a C++0x “Uniform Random Number Engine” and can be used in any context where such an object is expected.

Note that wrapping a counter based RNG with a traditional API in this way obscures much of the power of counter based PRNGs. Nevertheless, it may be of value in applications that are already coded to work with the C++0x random number engines.

The MicroURNG template in MicroURNG.hpp

provides the more limited functionality of a C++0x “Uniform

Random Number Generator”, but leaves the application in control of counters and keys and hence may be preferable to the

Engine template. For example, a MicroURNG

allows one to use C++0x “Random Number

Distributions” without giving up control over the counters and keys.

Public Types

typedef CBRNG cbrng_type

typedef CBRNG::ctr_type ctr_type

typedef CBRNG::key_type key_type

typedef CBRNG::ukey_type ukey_type

typedef ctr_type::value_type result_type

typedef size_t elem_type

Public Functions

inline explicit Engine()

inline explicit Engine(result_type r)

inline Engine(Engine &e)

inline Engine(const Engine &e)

template<typename SeedSeq>
inline explicit Engine(SeedSeq &s)

inline void seed(result_type r)

template<typename SeedSeq>
inline void seed(SeedSeq &s)

inline void seed()

inline result_type operator()()

inline void discard(R123_ULONG_LONG skip)

inline explicit Engine(const ukey_type &uk)

inline explicit Engine(ukey_type &uk)

inline void seed(const ukey_type &uk)

inline void seed(ukey_type &uk)

inline ctr_type operator()(const ctr_type &c) const

inline ukey_type getseed() const

inline std::pair<ctr_type, elem_type> getcounter() const

inline void setcounter(const ctr_type &_c, elem_type _elem)

Public Static Functions

static inline R123_CONSTEXPR result_type min R123_NO_MACRO_SUBST ()

static inline R123_CONSTEXPR result_type max R123_NO_MACRO_SUBST ()

Public Static Attributes

static const result_type _Min = 0

static const result_type _Max = ~((result_type)0)

Protected Functions

inline void fix_invariant()

Protected Attributes

cbrng_type b

key_type key

ukey_type ukey

ctr_type c

elem_type elem

ctr_type v

Friends

inline friend bool operator==(const Engine &lhs, const Engine &rhs)

inline friend bool operator!=(const Engine &lhs, const Engine &rhs)

inline friend std::ostream &operator<<(std::ostream &os, const Engine &be)

inline friend std::istream &operator>>(std::istream &is, Engine &be)

class EnqueueArgs

#include <opencl.hpp>

Public Functions

inline EnqueueArgs(NDRange global)

inline EnqueueArgs(NDRange global, NDRange local)

inline EnqueueArgs(NDRange offset, NDRange global, NDRange local)

inline EnqueueArgs(Event e, NDRange global)

inline EnqueueArgs(Event e, NDRange global, NDRange local)

inline EnqueueArgs(Event e, NDRange offset, NDRange global, NDRange local)

inline EnqueueArgs(const vector<Event> &events, NDRange global)

inline EnqueueArgs(const vector<Event> &events, NDRange global, NDRange local)

inline EnqueueArgs(const vector<Event> &events, NDRange offset, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, NDRange global)

inline EnqueueArgs(CommandQueue &queue, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, NDRange offset, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, Event e, NDRange global)

inline EnqueueArgs(CommandQueue &queue, Event e, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, Event e, NDRange offset, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, const vector<Event> &events, NDRange global)

inline EnqueueArgs(CommandQueue &queue, const vector<Event> &events, NDRange global, NDRange local)

inline EnqueueArgs(CommandQueue &queue, const vector<Event> &events, NDRange offset, NDRange global, NDRange local)

Private Members

CommandQueue queue_

const NDRange offset_

const NDRange global_

const NDRange local_

vector<Event> events_

Friends

friend class KernelFunctor

class Event : public cl::detail::Wrapper<cl_event>

#include <opencl.hpp>

Class interface for cl_event.

See also

cl_event

Note

Copies of these objects are shallow, meaning that the copy will refer to the same underlying cl_event as the original. For details, see clRetainEvent() and clReleaseEvent().

Subclassed by cl::UserEvent

Public Functions

inline Event()

Default constructor - initializes to NULL.

inline explicit Event(const cl_event &event, bool retainObject = false)

Constructor from cl_event - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. This effectively transfers ownership of a refcount on the cl_event into the new Event object.

inline Event &operator=(const cl_event &rhs)

Assignment operator from cl_event - takes ownership.

This effectively transfers ownership of a refcount on the rhs and calls clReleaseEvent() on the value previously held by this instance.

template<typename T>
inline cl_int getInfo(cl_event_info name, T *param) const

Wrapper for clGetEventInfo().

template<cl_event_info name>
inline detail::param_traits<detail::cl_event_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetEventInfo() that returns by value.

template<typename T>
inline cl_int getProfilingInfo(cl_profiling_info name, T *param) const

Wrapper for clGetEventProfilingInfo().

template<cl_profiling_info name>
inline detail::param_traits<detail::cl_profiling_info, name>::param_type getProfilingInfo(cl_int *err = NULL) const

Wrapper for clGetEventProfilingInfo() that returns by value.

inline cl_int wait() const

Blocks the calling thread until this event completes.

Wraps clWaitForEvents().

inline cl_int setCallback (cl_int type, void(CL_CALLBACK *pfn_notify)(cl_event, cl_int, void *), void *user_data=NULL)

Registers a user callback function for a specific command execution status.

Wraps clSetEventCallback().

Public Static Functions

static inline cl_int waitForEvents(const vector<Event> &events)

Blocks the calling thread until every event specified is complete.

Wraps clWaitForEvents().

struct f2_params_t

Public Members

const I3CLSimFunction *unbiasedSpectrum

const I3CLSimFunction *wavelengthGenerationBias

struct f_params_t

Public Members

const I3CLSimFunction *phaseRefIndex

const I3CLSimFunction *wavelengthGenerationBias

class GenerateStepPreCalculator

Public Functions

GenerateStepPreCalculator(I3RandomServicePtr randomService, double angularDist_a = 0.39, double angularDist_b = 2.61, std::size_t numberOfValues = 102400)

~GenerateStepPreCalculator()

inline void GetAngularCosSinValue(double &angular_cos, double &angular_sin, double &random_value)

Private Types

typedef std::vector<std::pair<std::pair<double, double>, double>> queueVector_t

Private Functions

void FeederThread(unsigned int threadId, uint64_t initialRngState, uint32_t rngA)

void RegenerateValues()

Private Members

double one_over_angularDist_a_

double angularDist_b_

double angularDist_I_

std::size_t numberOfValues_

std::size_t index_

boost::shared_ptr<queueVector_t> currentVector_

I3CLSimQueue<boost::shared_ptr<queueVector_t>> queueFromFeederThreads_

std::vector<boost::shared_ptr<boost::thread>> feederThreads_

template<typename Func, typename Arg0>
struct GetInfoFunctor0

#include <opencl.hpp>

Public Functions

inline cl_int operator()(cl_uint param, size_type size, void *value, size_type *size_ret)

Public Members

Func f_

const Arg0 &arg0_

template<typename Func, typename Arg0, typename Arg1>
struct GetInfoFunctor1

#include <opencl.hpp>

Public Functions

inline cl_int operator()(cl_uint param, size_type size, void *value, size_type *size_ret)

Public Members

Func f_

const Arg0 &arg0_

const Arg1 &arg1_

class I3CLSimCUDADevice

#include <I3CLSimCUDADevice.h>

Describes an OpenCL platform, device and a few device-specific parameters.

Public Functions

std::string GetDeviceName() const

inline bool IsCPU() const

inline bool IsGPU() const

std::size_t GetMaxComputeUnits() const

std::size_t GetMaxWorkItemSize() const

std::size_t GetMaxWorkGroupSize() const

std::size_t GetMaxClockFrequencyMhz() const

std::size_t GetGlobalMemSize() const

std::size_t GetMaxConstantBufferSize() const

std::size_t GetLocalMemSize() const

bool HasDedicatedLocalMem() const

bool HasErrorCorrectionSupport() const

bool IsAvailable() const

std::string GetVendor() const

inline int GetDeviceNumber() const

Public Static Functions

static boost::shared_ptr<std::vector<I3CLSimCUDADevice>> GetAllDevices()

Private Functions

I3CLSimCUDADevice(int number)

I3CLSimCUDADevice() = delete

Private Members

int deviceNumber_

boost::shared_ptr<cudaDeviceProp> device_

Friends

friend bool operator==(const I3CLSimCUDADevice&, const I3CLSimCUDADevice&)

class I3CLSimLightSourcePropagatorGeant4 : public I3LightSourcePropagator

#include <I3CLSimLightSourcePropagatorGeant4.h>

Public Functions

I3CLSimLightSourcePropagatorGeant4(std::string physicsListName = default_physicsListName, double maxBetaChangePerStep = default_maxBetaChangePerStep, uint32_t maxNumPhotonsPerStep = default_maxNumPhotonsPerStep, bool collectParticleHistory = false)

virtual ~I3CLSimLightSourcePropagatorGeant4()

virtual void SetRandomService(I3RandomServicePtr random)

virtual void SetWlenBias(I3CLSimFunctionConstPtr wlenBias)

Sets the wavelength bias. Set this to a constant value of 1 if you do not need biased photon generation. The Cherenkov spectrum will be multiplied by this value at each wavelength. This will influence the number of photons produced. Will throw if used after the call to Initialize().

virtual void SetMediumProperties(I3CLSimMediumPropertiesConstPtr mediumProperties)

Sets the medium properties. Will throw if used after the call to Initialize().

virtual void Initialize()

Initializes the simulation. Will throw if already initialized.

virtual bool IsInitialized() const

Returns true if initialized. Never throws.

inline virtual bool IsValidForLightSource(const I3LightSource &source)

virtual I3MCTreePtr Convert(I3LightSourceConstPtr&, uint32_t, secondary_callback, step_callback)

Public Static Attributes

static const std::string default_physicsListName = "QGSP_BERT_EMV"

static const double default_maxBetaChangePerStep = 10. * I3Units::perCent

static const uint32_t default_maxNumPhotonsPerStep = 200

Private Functions

void LogGeant4Messages(bool allAsWarn = false) const

SET_LOGGER ("I3CLSimLightSourcePropagatorGeant4")

Private Members

mutable boost::shared_ptr<I3CLSimQueue<boost::shared_ptr<std::pair<const std::string, bool>>>> queueFromGeant4Messages_

I3RandomServicePtr randomService_

uint32_t randomSeed_

std::string physicsListName_

double maxBetaChangePerStep_

uint32_t maxNumPhotonsPerStep_

bool initialized_

bool collectParticleHistory_

std::unique_ptr<G4RunManager> runManager_

I3CLSimFunctionConstPtr wlenBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

Private Static Attributes

static std::atomic<bool> thereCanBeOnlyOneGeant4

struct I3CLSimLightSourceToStepConverterFlasher : public I3LightSourceToStepConverter

#include <I3CLSimLightSourceToStepConverterFlasher.h>

A particle-to-step converter for IceCube LED flashers.

Public Functions

I3CLSimLightSourceToStepConverterFlasher(I3CLSimFunctionConstPtr flasherSpectrumNoBias, I3CLSimSpectrumTablePtr spectrumTable, I3CLSimRandomValueConstPtr angularProfileDistributionPolar, I3CLSimRandomValueConstPtr angularProfileDistributionAzimuthal, I3CLSimRandomValueConstPtr timeDelayDistribution, bool interpretAngularDistributionsInPolarCoordinates = default_interpretAngularDistributionsInPolarCoordinates, uint32_t photonsPerStep = default_photonsPerStep)

Initializes a new converter object for a specific flasher type.

flasherSpectrumNoBias: the flasher wavelength spectrum without any bias factors for DOM acceptance.

spectrumTable: the global spectrum table used for OpenCL code generation.

angularProfileDistributionPolar: a random number distribution of polar angles w.r.t. to the flasher orientation used for smearing of initial photon directions. The distribution is assumed to take a single run-time parameter: the “width” as specified in I3FlasherPulse.

angularProfileDistributionAzimuthal: a random number distribution of azimuthal angles w.r.t. to the flasher orientation used for smearing of initial photon directions. The distribution is assumed to take a single run-time parameter: the “width” as specified in I3FlasherPulse.

timeDelayDistribution: a distribution of time delays w.r.t. the nominal flasher time. Values from this distribution will be added to the nominal flasher time to generate photon starting times. The distribution is assumed to take a single run-time parameter: the “width” as specified in I3FlasherPulse.

interpretAngularDistributionsInPolarCoordinates: interpret the angularProfileDistributionPolar and angularProfileDistributionAzimuthal distributions in polar coordinates. Instead of a shift in the polar and azimuthal directions, the polar shift is a shift perpendicular from the current direction. The azimuthal shift is a rotation around the old axis. Since the origin of this rotation is undefined, it is best to use a uniform range of angles from 0deg to 360deg for the azimuthal shift in this case.

photonsPerStep: instead of creating photons directly, this class generates so-called steps (bunches of photons described by a single object). This controls how many photons there will be per step (some steps may have less photons than specified here to accomodate all possible numbers of photons). All random distributions are only sampled once per step, not once per photon, thus having too many photons per step may introduce too much grnaularity. This should not be a problem for flashers since they generate a large number of photons anyway.

virtual ~I3CLSimLightSourceToStepConverterFlasher()

virtual void SetBunchSizeGranularity(uint64_t num)

virtual void SetMaxBunchSize(uint64_t num)

virtual void SetRandomService(I3RandomServicePtr random)

virtual void SetWlenBias(I3CLSimFunctionConstPtr wlenBias)

virtual void SetMediumProperties(I3CLSimMediumPropertiesConstPtr mediumProperties)

virtual void Initialize()

virtual bool IsInitialized() const

virtual void EnqueueLightSource(const I3LightSource &lightSource, uint32_t identifier)

virtual void EnqueueBarrier()

virtual bool BarrierActive() const

virtual bool MoreStepsAvailable() const

virtual I3SimStepSeriesConstPtr GetConversionResultWithBarrierInfo(bool &barrierWasReset, double timeout = NAN)

Public Static Attributes

static const uint32_t default_photonsPerStep = 400

static const bool default_interpretAngularDistributionsInPolarCoordinates = false

Private Functions

I3SimStepSeriesConstPtr MakeSteps(bool &barrierWasReset)

void FillStep(I3SimStep &step, uint32_t numberOfPhotons, const I3FlasherPulse &flasherPulse, uint32_t identifier)

Private Members

std::deque<LightSourceData_t> inputQueue_

I3RandomServicePtr randomService_

bool initialized_

bool barrier_is_enqueued_

uint64_t bunchSizeGranularity_

uint64_t maxBunchSize_

uint32_t photonsPerStep_

uint8_t spectrumSourceTypeIndex_

I3CLSimFunctionConstPtr flasherSpectrumNoBias_

I3CLSimFunctionConstPtr wlenBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

double photonNumberCorrectionFactorForBias_

I3CLSimRandomValueConstPtr angularProfileDistributionPolar_

I3CLSimRandomValueConstPtr angularProfileDistributionAzimuthal_

I3CLSimRandomValueConstPtr timeDelayDistribution_

bool interpretAngularDistributionsInPolarCoordinates_

struct I3CLSimLightSourceToStepConverterPPC : public I3LightSourceToStepConverter

#include <I3CLSimLightSourceToStepConverterPPC.h>

A particle-to-step converter for cascades using pre-defined parameterizations.

Public Functions

I3CLSimLightSourceToStepConverterPPC(uint32_t photonsPerStep = default_photonsPerStep, uint32_t highPhotonsPerStep = default_highPhotonsPerStep, double useHighPhotonsPerStepStartingFromNumPhotons = default_useHighPhotonsPerStepStartingFromNumPhotons)

virtual ~I3CLSimLightSourceToStepConverterPPC()

inline void SetUseCascadeExtension(bool v)

virtual void SetBunchSizeGranularity(uint64_t num)

virtual void SetMaxBunchSize(uint64_t num)

virtual void SetRandomService(I3RandomServicePtr random)

virtual void SetWlenBias(I3CLSimFunctionConstPtr wlenBias)

virtual void SetMediumProperties(I3CLSimMediumPropertiesConstPtr mediumProperties)

virtual void Initialize()

virtual bool IsInitialized() const

virtual void EnqueueLightSource(const I3LightSource &lightSource, uint32_t identifier)

virtual void EnqueueBarrier()

virtual bool BarrierActive() const

virtual bool MoreStepsAvailable() const

virtual I3SimStepSeriesConstPtr GetConversionResultWithBarrierInfo(bool &barrierWasReset, double timeout = NAN)

Public Static Attributes

static const uint32_t default_photonsPerStep = 200

static const uint32_t default_highPhotonsPerStep = 0

static const double default_useHighPhotonsPerStepStartingFromNumPhotons = 1.0e9

Private Types

typedef boost::variant<CascadeStepData_t, MuonStepData_t, BarrierData_t> StepData_t

Private Functions

I3SimStepSeriesConstPtr MakeSteps(bool &barrierWasReset)

Private Members

std::deque<StepData_t> stepGenerationQueue_

I3RandomServicePtr randomService_

uint64_t rngState_

uint32_t rngA_

bool initialized_

bool barrier_is_enqueued_

uint64_t bunchSizeGranularity_

uint64_t maxBunchSize_

uint32_t photonsPerStep_

uint32_t highPhotonsPerStep_

double useHighPhotonsPerStepStartingFromNumPhotons_

bool useCascadeExtension_

I3CLSimFunctionConstPtr wlenBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

std::vector<double> meanPhotonsPerMeterInLayer_

boost::shared_ptr<GenerateStepPreCalculator> preCalc_

Private Static Functions

static void GenerateStep(I3SimStep &newStep, const I3Particle &p, double particleDir_x, double particleDir_y, double particleDir_z, uint32_t identifier, uint32_t photonsPerStep, const double &longitudinalPos, GenerateStepPreCalculator &preCalc)

static void GenerateStepForMuon(I3SimStep &newStep, const I3Particle &p, double particleDir_x, double particleDir_y, double particleDir_z, uint32_t identifier, uint32_t photonsPerStep, double length)

template<typename OutputMapType>
class I3CLSimModule : public I3ConditionalModule

#include <I3CLSimModule.h>

Public Functions

I3CLSimModule(const I3Context &ctx)

Builds an instance of this class

virtual ~I3CLSimModule()

Destroys an instance of this class

virtual void Configure()

This module takes a configuration parameter and so it must be configured.

virtual void Process()

The module needs full control of all streams -> implement Process()!

virtual void Finish()

Warn the user if the module is aborted prematurely

virtual bool ShouldDoProcess(I3FramePtr frame)

Hack to allow buffering. This ShouldDoProcess overrides the I3ConditionalModule implementation and always returns true. The original ShouldDoProcess() is used in our Process() in order mark frames in case they should not be touched. They will be put in the buffer in any case, however.

This should retain frame ordering.

Private Functions

bool DigestOtherFrame(I3FramePtr frame, bool startThread = true)

The module needs to process Physics frames

void DigestGeometry(I3FramePtr frame)

The module needs to process Geometry frames

double GetLightSourceEnergy(I3FramePtr frame)

Getting energy from light source to make sure to process the right number of frames

I3CLSimModule()

I3CLSimModule(const I3CLSimModule&)

I3CLSimModule &operator=(const I3CLSimModule&)

void StopThread()

void StartThread()

void Thread_starter()

bool Thread(boost::this_thread::disable_interruption &di)

This thread takes care of passing steps from Geant4 to OpenCL

std::size_t FlushFrameCache()

void ConvertMCTreeToLightSources(const I3MCTree &mcTree, std::deque<I3LightSource> &lightSources, std::deque<double> &timeOffsets)

void ConvertFlasherPulsesToLightSources(const I3FlasherPulseSeries &flasherPulses, std::deque<I3LightSource> &lightSources, std::deque<double> &timeOffsets)

SET_LOGGER ("I3CLSimModule")

Private Members

std::vector<I3Frame::Stream> workOnTheseStops_

Parameter: work on MCTrees found in the stream types (“stops”) specified in this list.

std::set<I3Frame::Stream> workOnTheseStops_set_

I3LightSourceParameterizationSeries parameterizationList_

Parameter: An instance I3LightSourceParameterizationSeries specifying the fast simulation parameterizations to be used.

I3RandomServicePtr randomService_

Parameter: A random number generating service (derived from I3RandomService).

bool generateCherenkovPhotonsWithoutDispersion_

Parameter: The wavelength of the generated Cherenkov photons will be generated according to a spectrum without dispersion. This does not change the total number of photons, only the distribution of wavelengths.

I3CLSimFunctionConstPtr wavelengthGenerationBias_

Parameter: An instance of I3CLSimFunction describing the reciprocal weight a photon gets assigned as a function of its wavelength. You can set this to the wavelength depended acceptance of your DOM to pre-scale the number of generated photons.

I3CLSimMediumPropertiesConstPtr mediumProperties_

Parameter: An instance of I3CLSimMediumProperties describing the ice/water properties.

I3CLSimSpectrumTableConstPtr spectrumTable_

Parameter: All spectra that could be requested by an I3SimStep. If set to NULL/None, only spectrum #0 (Cherenkov photons) will be available.

unsigned int maxNumParallelEvents_

Parameter: Maximum number of events that will be processed by the GPU in parallel.

unsigned int maxNumParallelEventsSecondFlush_

double totalEnergyToProcess_

Parameter: Maximum energy to that will be processed by the GPU in parallel.

I3CLSimOpenCLDeviceSeries openCLDeviceList_

Parameter: A vector of I3CLSimOpenCLDevice objects, describing the devices to be used for simulation.

double DOMRadius_

Parameter: The DOM radius used during photon tracking.

double DOMOversizeFactor_

Parameter: Specifiy the “oversize factor” (i.e. DOM radius scaling factor).

bool ignoreNonIceCubeOMNumbers_

Parameter: Ignore string numbers < 1 and OM numbers > 60. (AMANDA and IceTop)

std::string MCTreeName_

Parameter: Name of the I3MCTree frame object. All particles except neutrinos will be read from this tree.

std::string flasherPulseSeriesName_

Parameter: Name of the I3FlasherPulseSeries frame object. Flasher pulses will be read from this object. Set this to the empty string to disable flashers.

std::string photonSeriesMapName_

Parameter: Name of the I3CLSimPhotonSeriesMap frame object that will be written to the frame.

std::string omKeyMaskName_

Parameter: Name of a I3VectorOMKey with masked OMKeys. DOMs in this list will not record I3Photons.

bool ignoreMuons_

Parameter: If set to True, muons will not be propagated.

std::string geant4PhysicsListName_

Parameter: Geant4 physics list name. Examples are “QGSP_BERT_EMV” and “QGSP_BERT”.

double geant4MaxBetaChangePerStep_

Parameter: Maximum change of beta=v/c per Geant4 step.

uint32_t geant4MaxNumPhotonsPerStep_

Parameter: Approximate maximum number of Cherenkov photons generated per step by Geant4.

std::string statisticsName_

Parameter: Collect statistics in this frame object (e.g. number of photons generated or reaching the DOMs)

bool collectStatistics_

std::vector<int> ignoreStrings_

Parameter: Ignore all OMKeys with these string IDs.

std::vector<unsigned int> ignoreDomIDs_

Parameter: Ignore all OMKeys with these DOM IDs.

std::vector<std::string> ignoreSubdetectors_

Parameter: Ignore all OMKeys with these subdetector names.

bool splitGeometryIntoPartsAcordingToPosition_

Parmeter: If you have a geometry with multiple OMs per floor (e.g. Antares or KM3NeT-tower-like), it will internally get split into subdetectors to save memory on the GPU. This is 100% transparent. By default the split is according to the “floor index” of an OM on a floor. If you enable this option, the split will also be done according to the x-y projected positions of the OMs per string. This may be necessary for “tower” geometries.

bool useHardcodedDeepCoreSubdetector_

Parameter: Split off DeepCore as its own two subdetectors (upper and lower part). This may save constant memory on your GPU. Assumes that strings [79..86] are DeepCore strings with an upper part at z>-30m and a lower part at z<-30m.

bool enableDoubleBuffering_

Parmeter: Disables or enables double-buffered GPU usage. Double buffering will use two command queues and two sets of input and output buffers in order to transfer data to the GPU while a kernel is executing on the other buffer. This has been observed to yield empty results results on older drivers for the nVidia architecture, so it is disabled by default.

Before enabling this for a certain driver/hardware combination, make sure that both correct results are returned. Most of the time the second buffer results are always empty, so this error should be easy to observe.

bool doublePrecision_

Parmeter: Enables double-precision support in the kernel. This slows down calculations and requires more memory. The performance hit is minimal on CPUs but up to an order of magnitude on GPUs.

bool stopDetectedPhotons_

Parmeter: Configures behaviour for photons that hit a DOM. If this is true (the default) photons will be stopped once they hit a DOM. If this is false, they continue to propagate.

bool saveAllPhotons_

Parmeter: Saves all photons, even if they don’t hit a DOM. Cannot be used with “StopDetectedPhotons”.

double saveAllPhotonsPrescale_

Parmeter: Sets the prescale factor of photons being generated in “saveAllPhotons” mode. Only this fraction of photons is actually generated.

double fixedNumberOfAbsorptionLengths_

Parameter: Sets the number of absorption lengths each photon should be propagated. If set to NaN (the default), the number is sampled from an exponential distribution. (This is what you want for “normal” propagation.)

Use this override for table-making.

double pancakeFactor_

Parameter: Sets the “pancake” factor for DOMs. For standard oversized-DOM simulations, this should be the radius oversizing factor. This will flatten the DOM in the direction parallel to the photon. The DOM will have a pancake-like shape, elongated in the directions perpendicular to the photon direction.

The DOM radius (supplied by the geometry) must also include the oversizing factor.

uint32_t photonHistoryEntries_

Parmeter: Sets the number of scattering step positions that are saved for a photon hitting a DOM. The last N photons are saved if there are more scattering points than available entries.

uint32_t limitWorkgroupSize_

Parmeter: Limits the OpenCL workgroup size (the number of bunches to be processed in parallel). If set to zero (the default) the largest possible workgroup size will be chosen.

double closestDOMDistanceCutoff_

Parameter: do not even start light from sources that do not have any DOMs closer to to them than this distance. (default is 300m)

boost::shared_ptr<I3Surfaces::ExtrudedPolygon> detectorHull_

boost::shared_ptr<boost::thread> threadObj_

boost::condition_variable_any threadStarted_cond_

boost::mutex threadStarted_mutex_

bool threadStarted_

bool threadFinishedOK_

std::vector<uint64_t> numBunchesSentToOpenCL_

std::map<uint32_t, uint64_t> photonNumGeneratedPerParticle_

std::map<uint32_t, double> photonWeightSumGeneratedPerParticle_

bool geometryIsConfigured_

uint32_t currentParticleCacheIndex_

double totalSimulatedEnergyForFlush_

double totalSimulatedEnergy_

double totalSimulatedEnergySecondFlush_

uint64_t totalNumParticlesForFlush_

std::vector<I3CLSimRandomValueConstPtr> wavelengthGenerators_

I3CLSimSimpleGeometryFromI3GeometryPtr geometry_

std::vector<I3CLSimStepToPhotonConverterOpenCLPtr> openCLStepsToPhotonsConverters_

I3CLSimLightSourceToStepConverterGeant4Ptr geant4ParticleToStepsConverter_

std::size_t frameListPhysicsFrameCounter_

std::vector<I3FramePtr> frameList_

std::deque<I3FramePtr> frameList2_

std::vector<boost::shared_ptr<OutputMapType>> photonsForFrameList_

std::vector<int32_t> currentPhotonIdForFrame_

std::vector<bool> frameIsBeingWorkedOn_

std::vector<std::set<ModuleKey>> maskedOMKeys_

std::map<uint32_t, particleCacheEntry> particleCache_

class I3CLSimOpenCLDevice

#include <I3CLSimOpenCLDevice.h>

Describes an OpenCL platform, device and a few device-specific parameters.

Public Functions

~I3CLSimOpenCLDevice()

I3CLSimOpenCLDevice(const std::string &platformName, const std::string &deviceName)

I3CLSimOpenCLDevice(const std::string &platformName, const std::string &deviceName, bool useNativeMath, uint32_t approximateNumberOfWorkItems)

boost::shared_ptr<std::vector<I3CLSimOpenCLDevice>> SplitDevice() const

inline const std::string &GetPlatformName() const

inline const std::string &GetDeviceName() const

inline void SetUseNativeMath(bool useit)

inline bool GetUseNativeMath() const

inline void SetApproximateNumberOfWorkItems(uint32_t newnumber)

inline uint32_t GetApproximateNumberOfWorkItems() const

bool IsCPU() const

bool IsGPU() const

std::size_t GetMaxComputeUnits() const

std::size_t GetMaxWorkItemSize() const

std::size_t GetMaxWorkGroupSize() const

std::size_t GetMaxClockFrequencyMhz() const

std::size_t GetGlobalMemSize() const

std::size_t GetMaxConstantBufferSize() const

std::size_t GetLocalMemSize() const

bool HasDedicatedLocalMem() const

bool HasErrorCorrectionSupport() const

bool IsAvailable() const

std::string GetVendor() const

std::string GetDriverVersion() const

std::string GetDeviceVersion() const

std::string GetExtensions() const

inline boost::shared_ptr<cl::Platform> GetPlatformHandle() const

inline boost::shared_ptr<cl::Device> GetDeviceHandle() const

Public Static Functions

static boost::shared_ptr<std::vector<I3CLSimOpenCLDevice>> GetAllDevices()

Private Functions

I3CLSimOpenCLDevice()

void DoInit(const std::string &platformName, const std::string &deviceName)

Private Members

boost::shared_ptr<cl::Platform> platform_

boost::shared_ptr<cl::Device> device_

std::string platformName_

std::string deviceName_

bool useNativeMath_

uint32_t approximateNumberOfWorkItems_

Private Static Functions

static void InitializeStaticStuff()

Private Static Attributes

static bool lists_initialized_ = false

static std::vector<I3CLSimOpenCLDevice> allDevices_

Friends

friend bool operator==(const I3CLSimOpenCLDevice&, const I3CLSimOpenCLDevice&)

struct I3CLSimSpectrumTable

#include <I3CLSimSpectrumTable.h>

Stores all spectra that may be used during photon generation. If a spectrum that is being added should already exist, it is not added a second time. The previous version is re-used.

Spectrum number “0” is always the Cherenkov spectrum and does not need to be added by the user.

Public Functions

I3CLSimSpectrumTable()

~I3CLSimSpectrumTable()

inline const std::vector<I3CLSimFunctionConstPtr> &GetSpectra() const

Returns the vector of spectra. The first spectrum is the Cherenkov spectrum and will be set to NULL.

std::size_t append(I3CLSimFunctionConstPtr newSpectrum)

Add a new spectrum. The spectrum index will be returned. This can also be used to retrieve the index of an existing spectrum.

inline I3CLSimFunctionConstPtr operator[](std::size_t index) const

inline std::size_t size() const

Private Members

std::vector<I3CLSimFunctionConstPtr> spectra_

struct I3CLSimStepToPhotonConverterCUDA : public I3StepToPhotonConverter

#include <I3CLSimStepToPhotonConverterCUDA.h>

Creates photons from a given list of steps and propagates them to a DOM using a CUDA program.

Public Functions

I3CLSimStepToPhotonConverterCUDA(I3RandomServicePtr randomService, bool useNativeMath = default_useNativeMath)

virtual ~I3CLSimStepToPhotonConverterCUDA()

void SetWorkgroupSize(std::size_t val)

Sets the workgroup size. A value of 0 uses the maximum possible workgroup size for the kernel.

Will throw if already initialized.

void SetMaxNumWorkitems(std::size_t val)

Sets the number of parallel work items.

Will throw if already initialized.

std::size_t GetWorkgroupSize() const

Gets the current workgroup size.

std::size_t GetMaxNumWorkitems() const

Gets the number of parallel work items.

void SetDevice(const I3CLSimCUDADevice &device)

Sets the CUDA device.

Will throw if already initialized.

uint64_t GetMaxWorkgroupSize() const

Returns the maximum workgroup size for the current kernel.

Will throw if compilation fails. Will throw if initialized.

void SetEnableDoubleBuffering(bool value)

Disables or enables double-buffered GPU usage. Double buffering will use two command queues and two sets of input and output buffers in order to transfer data to the GPU while a kernel is executing on the other buffer.

This has been observed to yield empty results results on older drivers for the nVidia architecture, so it is disabled by default.

Before enabling this for a certain driver/hardware combination, make sure that both correct results are returned. Most of the time the second buffer results are always empty, so this error should be easy to observe.

Will throw if already initialized.

bool GetEnableDoubleBuffering() const

Returns true if double buffering is enabled.

void SetDoublePrecision(bool value)

Enables double-precision support in the kernel. This slows down calculations and requires more memory.

The performance hit is minimal on CPUs but up to an order of magnitude on GPUs.

Will throw if already initialized.

bool GetDoublePrecision() const

Returns true if double precision is enabled.

void SetDOMPancakeFactor(double value)

Sets the “pancake” factor for DOMs. For standard oversized-DOM simulations, this should be the radius oversizing factor. This will flatten the DOM in the direction parallel to the photon. The DOM will have a pancake-like shape, elongated in the directions perpendicular to the photon direction.

The DOM radius (supplied by the geometry) must also include the oversizing factor.

Will throw if already initialized.

double GetDOMPancakeFactor() const

Returns the “pancake” factor for DOMs.

virtual void SetWlenGenerators(const std::vector<I3CLSimRandomValueConstPtr> &wlenGenerators)

Sets the wavelength generators. The first generator (index 0) is assumed to return a Cherenkov spectrum that may have a bias applied to it. This bias factor needs to be set using SetWlenBias(). All other generator indices are assumed to be for flasher/laser light generation. During generation, no Cherenkov angle rotation will be applied to those photons with indices >= 1. Will throw if used after the call to Initialize().

virtual void SetWlenBias(I3CLSimFunctionConstPtr wlenBias)

Sets the wavelength weights. Set this to a constant value of 1 if you do not need biased photon generation. The wavelength spectrum set with SetWlenGenerator() is assumed to have a biassing factor already applied to it. This call sets this factor in order to be able to assign correct weights. Will throw if used after the call to Initialize().

virtual void SetMediumProperties(I3CLSimMediumPropertiesConstPtr mediumProperties)

Sets the medium properties. Will throw if used after the call to Initialize().

virtual void SetGeometry(I3SimpleGeometryConstPtr geometry)

Sets the geometry. Will throw if used after the call to Initialize().

virtual void Initialize()

Initializes the simulation. Will throw if already initialized.

virtual bool IsInitialized() const

Returns true if initialized. Never throws.

virtual void EnqueueSteps(I3SimStepSeriesConstPtr steps, uint32_t identifier)

Adds a new I3SimStepSeries to the queue. The resulting I3CLSimPhotonSeries can be retrieved from the I3StepToPhotonConverter after some processing time.

Enqueuing a vector after calling EnqueueBarrier will throw if not all photons have been retrieved.

Will throw if not initialized.

virtual std::size_t QueueSize() const

Reports the current queue size. The queue works asynchronously, so this value will probably have changed once you use it.

Will throw if not initialized.

virtual bool MorePhotonsAvailable() const

Returns true if more photons are available. If the return value is false, the current simulation is finished and a new step vector may be set.

Will throw if not initialized.

virtual I3StepToPhotonConverter::ConversionResult_t GetConversionResult()

Returns a bunch of photons stored in a vector<I3CLSimPhoton>.

The return value is a pair<uint, vector<I3CLSimPhoton> >. The integer is the same identifier as specified in the call to EnqueueSteps().

Might block if no photons are available.

Will throw if not initialized.

virtual std::map<std::string, double> GetStatistics() const

inline double GetTotalDeviceTime() const

inline double GetTotalHostTime() const

inline double GetTotalQueueTime() const

inline uint64_t GetNumKernelCalls() const

inline uint64_t GetTotalNumPhotonsGenerated() const

inline uint64_t GetTotalNumPhotonsAtDOMs() const

Public Static Attributes

static const bool default_useNativeMath = true

Private Types

typedef std::pair<uint32_t, I3SimStepSeriesConstPtr> ToOpenCLPair_t

Private Functions

void ServiceThread(unsigned, boost::shared_ptr<boost::barrier>&)

void ServiceThread_impl(unsigned, boost::shared_ptr<boost::barrier>)

SET_LOGGER ("I3CLSimStepToPhotonConverterCUDA")

Private Members

statistics_bundle statistics_

std::vector<boost::thread> threads_

boost::shared_ptr<I3CLSimQueue<ToOpenCLPair_t>> inputQueue_

boost::shared_ptr<I3CLSimQueue<I3StepToPhotonConverter::ConversionResult_t>> outputQueue_

I3RandomServicePtr randomService_

bool initialized_

bool compiled_

I3CLSimRandomValueInterpolatedDistributionConstPtr wlenGenerator_

I3CLSimFunctionFromTableConstPtr wlenBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

I3SimpleGeometryConstPtr geometry_

boost::shared_ptr<I3CLSimCUDADevice> device_

bool useNativeMath_

bool deviceIsSelected_

bool disableDoubleBuffering_

bool doublePrecision_

double pancakeFactor_

uint64_t maxWorkgroupSize_

std::size_t workgroupSize_

std::size_t maxNumWorkitems_

std::vector<uint64_t> MWC_RNG_x

std::vector<uint32_t> MWC_RNG_a

uint32_t maxNumOutputPhotons_

struct I3CLSimStepToPhotonConverterOpenCL : public I3StepToPhotonConverter

#include <I3CLSimStepToPhotonConverterOpenCL.h>

Creates photons from a given list of steps and propagates them to a DOM using an OpenCL-enabled algorithm.

Public Functions

I3CLSimStepToPhotonConverterOpenCL(I3RandomServicePtr randomService, bool useNativeMath = default_useNativeMath)

virtual ~I3CLSimStepToPhotonConverterOpenCL()

void SetWorkgroupSize(std::size_t val)

Sets the workgroup size. A value of 0 uses the maximum possible workgroup size for the kernel.

Will throw if already initialized.

void SetMaxNumWorkitems(std::size_t val)

Sets the number of parallel work items.

Will throw if already initialized.

std::size_t GetWorkgroupSize() const

Gets the current workgroup size.

std::size_t GetMaxNumWorkitems() const

Gets the number of parallel work items.

void SetMaxNumOutputPhotonsPerStep(std::uint32_t val)

void SetDevice(const I3CLSimOpenCLDevice &device)

Sets the OpenCL device.

Will throw if already initialized.

uint64_t GetMaxWorkgroupSize() const

Returns the maximum workgroup size for the current kernel.

Will throw if compilation fails. Will throw if initialized.

void SetEnableDoubleBuffering(bool value)

Disables or enables double-buffered GPU usage. Double buffering will use two command queues and two sets of input and output buffers in order to transfer data to the GPU while a kernel is executing on the other buffer.

This has been observed to yield empty results results on older drivers for the nVidia architecture, so it is disabled by default.

Before enabling this for a certain driver/hardware combination, make sure that both correct results are returned. Most of the time the second buffer results are always empty, so this error should be easy to observe.

Will throw if already initialized.

bool GetEnableDoubleBuffering() const

Returns true if double buffering is enabled.

void SetDoublePrecision(bool value)

Enables double-precision support in the kernel. This slows down calculations and requires more memory.

The performance hit is minimal on CPUs but up to an order of magnitude on GPUs.

Will throw if already initialized.

bool GetDoublePrecision() const

Returns true if double precision is enabled.

void SetStopDetectedPhotons(bool value)

Configures behaviour for photons that hit a DOM. If this is true (the default) photons will be stopped once they hit a DOM. If this is false, they continue to propagate.

Will throw if already initialized.

bool GetStopDetectedPhotons() const

Returns true if detected photons are stopped.

void SetSaveAllPhotons(bool value)

Tell the propagator to save all photons, regardless of detection. All photons will be assigned to DOM(0,0).

Will throw if already initialized.

bool GetSaveAllPhotons() const

Returns true if all photons are saved, regardless of detection.

void SetSaveAllPhotonsPrescale(double value)

Sets the prescale factor of photons being generated in “saveAllPhotons” mode. Only this fraction of photons is actually generated.

Will throw if already initialized.

double GetSaveAllPhotonsPrescale() const

Returns true if all photons are saved, regardless of detection.

void SetPhotonHistoryEntries(uint32_t value)

Sets the maximum number of entries in the photon history table. Each point in the table will store the position of the photon at each point of scatter. (Only the most recent points are stored if there are more scatters than available entries.)

Will throw if already initialized.

uint32_t GetPhotonHistoryEntries() const

Returns the maximum number of photon history entries.

void SetFixedNumberOfAbsorptionLengths(double value)

Sets the number of absorption lengths each photon should be propagated. If set to NaN (the default), the number is sampled from an exponential distribution. Use this override for table-making.

Will throw if already initialized.

double GetFixedNumberOfAbsorptionLengths() const

Returns number of absorption lengths each photon should be propagated. If set to NaN (the default), the number is sampled from an exponential distribution. Use this override for table-making.

void SetDOMPancakeFactor(double value)

Sets the “pancake” factor for DOMs. For standard oversized-DOM simulations, this should be the radius oversizing factor. This will flatten the DOM in the direction parallel to the photon. The DOM will have a pancake-like shape, elongated in the directions perpendicular to the photon direction.

The DOM radius (supplied by the geometry) must also include the oversizing factor.

Will throw if already initialized.

double GetDOMPancakeFactor() const

Returns the “pancake” factor for DOMs.

virtual void SetWlenGenerators(const std::vector<I3CLSimRandomValueConstPtr> &wlenGenerators)

Sets the wavelength generators. The first generator (index 0) is assumed to return a Cherenkov spectrum that may have a bias applied to it. This bias factor needs to be set using SetWlenBias(). All other generator indices are assumed to be for flasher/laser light generation. During generation, no Cherenkov angle rotation will be applied to those photons with indices >= 1. Will throw if used after the call to Initialize().

virtual void SetWlenBias(I3CLSimFunctionConstPtr wlenBias)

Sets the wavelength weights. Set this to a constant value of 1 if you do not need biased photon generation. The wavelength spectrum set with SetWlenGenerator() is assumed to have a biassing factor already applied to it. This call sets this factor in order to be able to assign correct weights. Will throw if used after the call to Initialize().

virtual void SetMediumProperties(I3CLSimMediumPropertiesConstPtr mediumProperties)

Sets the medium properties. Will throw if used after the call to Initialize().

virtual void SetGeometry(I3SimpleGeometryConstPtr geometry)

Sets the geometry. Will throw if used after the call to Initialize().

virtual void Compile()

Compiles the kernel. Can only be used after medium properties, geometry and device have been set.

Will throw if already initialized.

virtual std::string GetFullSource()

Gets the full generated OpenCL source code.

Will throw if not yet compiled.

virtual std::string GetPreambleSource()

Provide individual bits of the OpenCL source code.

std::string GetMediumPropertiesSource()

std::string GetWlenGeneratorSource()

std::string GetWlenBiasSource()

virtual std::string GetGeometrySource()

virtual std::string GetCollisionDetectionSource(bool header = true)

virtual void Initialize()

Initializes the simulation. Will throw if already initialized.

virtual bool IsInitialized() const

Returns true if initialized. Never throws.

virtual void EnqueueSteps(I3SimStepSeriesConstPtr steps, uint32_t identifier)

Adds a new I3SimStepSeries to the queue. The resulting I3CLSimPhotonSeries can be retrieved from the I3StepToPhotonConverter after some processing time.

Enqueuing a vector after calling EnqueueBarrier will throw if not all photons have been retrieved.

Will throw if not initialized.

virtual std::size_t QueueSize() const

Reports the current queue size. The queue works asynchronously, so this value will probably have changed once you use it.

Will throw if not initialized.

virtual bool MorePhotonsAvailable() const

Returns true if more photons are available. If the return value is false, the current simulation is finished and a new step vector may be set.

Will throw if not initialized.

virtual I3StepToPhotonConverter::ConversionResult_t GetConversionResult()

Returns a bunch of photons stored in a vector<I3CLSimPhoton>.

The return value is a pair<uint, vector<I3CLSimPhoton> >. The integer is the same identifier as specified in the call to EnqueueSteps().

Might block if no photons are available.

Will throw if not initialized.

virtual std::map<std::string, double> GetStatistics() const

inline double GetTotalDeviceTime() const

inline double GetTotalHostTime() const

inline double GetTotalQueueTime() const

inline uint64_t GetNumKernelCalls() const

inline uint64_t GetTotalNumPhotonsGenerated() const

inline uint64_t GetTotalNumPhotonsAtDOMs() const

Public Static Attributes

static const bool default_useNativeMath = true

Private Types

typedef std::pair<uint32_t, I3SimStepSeriesConstPtr> ToOpenCLPair_t

Private Functions

void SetupQueueAndKernel(const cl::Platform &platform, const cl::Device &device)

void OpenCLThread()

void OpenCLThread_impl(boost::this_thread::disable_interruption &di)

bool OpenCLThread_impl_uploadSteps(boost::this_thread::disable_interruption &di, const boost::posix_time::ptime &last_timestamp, bool &shouldBreak, unsigned int bufferIndex, uint32_t &out_stepsIdentifier, uint64_t &out_totalNumberOfPhotons, std::size_t &out_numberOfInputSteps, bool blocking = true)

void OpenCLThread_impl_downloadPhotons(boost::this_thread::disable_interruption &di, bool &shouldBreak, unsigned int bufferIndex, uint32_t stepsIdentifier)

void OpenCLThread_impl_runKernel(unsigned int bufferIndex, cl::Event &kernelFinishEvent, std::size_t numberOfInputSteps)

boost::posix_time::ptime DumpStatistics(const cl::Event &kernelFinishEvent, const boost::posix_time::ptime &last_timestamp, uint64_t totalNumberOfPhotons, bool starving, const std::string &platformName, const std::string &deviceName, uint64_t deviceProfilingResolution)

SET_LOGGER ("I3CLSimStepToPhotonConverterOpenCL")

Private Members

statistics_bundle statistics_

boost::shared_ptr<boost::thread> openCLThreadObj_

boost::condition_variable_any openCLStarted_cond_

boost::mutex openCLStarted_mutex_

bool openCLStarted_

boost::shared_ptr<I3CLSimQueue<ToOpenCLPair_t>> queueToOpenCL_

boost::shared_ptr<I3CLSimQueue<I3StepToPhotonConverter::ConversionResult_t>> queueFromOpenCL_

I3RandomServicePtr randomService_

bool initialized_

bool compiled_

std::vector<I3CLSimRandomValueConstPtr> wlenGenerators_

I3CLSimFunctionConstPtr wlenBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

I3SimpleGeometryConstPtr geometry_

I3CLSimOpenCLDevicePtr device_

bool useNativeMath_

bool deviceIsSelected_

bool disableDoubleBuffering_

bool doublePrecision_

bool stopDetectedPhotons_

bool saveAllPhotons_

double saveAllPhotonsPrescale_

double fixedNumberOfAbsorptionLengths_

double pancakeFactor_

uint32_t photonHistoryEntries_

std::string prependSource_

std::string mwcrngKernelSource_

std::string wlenGeneratorSource_

std::string wlenBiasSource_

std::string mediumPropertiesSource_

std::string geometrySource_

std::string propagationKernelSource_

std::vector<unsigned short> geoLayerToOMNumIndexPerStringSetInfo_

std::vector<int> stringIndexToStringIDBuffer_

std::vector<std::vector<unsigned int>> domIndexToDomIDBuffer_perStringIndex_

std::vector<boost::shared_ptr<cl::CommandQueue>> queue_

std::vector<boost::shared_ptr<cl::Kernel>> kernel_

boost::shared_ptr<cl::Context> context_

uint64_t maxWorkgroupSize_

std::size_t workgroupSize_

std::size_t maxNumWorkitems_

uint32_t maxNumOutputPhotonsPerStep_

std::vector<uint64_t> MWC_RNG_x

std::vector<uint32_t> MWC_RNG_a

boost::shared_ptr<cl::Buffer> deviceBuffer_MWC_RNG_x

boost::shared_ptr<cl::Buffer> deviceBuffer_MWC_RNG_a

std::vector<boost::shared_ptr<cl::Buffer>> deviceBuffer_InputSteps

std::vector<boost::shared_ptr<cl::Buffer>> deviceBuffer_OutputPhotons

std::vector<boost::shared_ptr<cl::Buffer>> deviceBuffer_CurrentNumOutputPhotons

std::vector<boost::shared_ptr<cl::Buffer>> deviceBuffer_PhotonHistory

boost::shared_ptr<cl::Buffer> deviceBuffer_GeoLayerToOMNumIndexPerStringSet

uint32_t maxNumOutputPhotons_

class I3CLSimStepToTableConverter : private noncopyable

#include <I3CLSimStepToTableConverter.h>

Public Functions

I3CLSimStepToTableConverter(I3CLSimOpenCLDevice device, clsim::tabulator::AxesConstPtr axes, size_t entriesPerStream, bool storeSquaredWeights, I3CLSimMediumPropertiesConstPtr medium, I3CLSimSpectrumTableConstPtr spectrumTable, double referenceArea, I3CLSimFunctionConstPtr wavelengthAcceptance, I3CLSimFunctionConstPtr angularAcceptance, I3RandomServicePtr rng)

virtual ~I3CLSimStepToTableConverter()

void EnqueueSteps(I3SimStepSeriesConstPtr, I3ParticleConstPtr)

void Finish()

inline size_t GetBunchSize() const

void WriteFITSFile(const std::string &fname, boost::python::dict tableHeader)

Private Types

typedef std::pair<I3SimStepSeriesConstPtr, I3ParticleConstPtr> bunch_t

Private Functions

void FetchSteps(cl::Kernel, I3RandomServicePtr)

float GetBinVolume(size_t i)

void Normalize()

SET_LOGGER ("I3CLSimStepToTableConverter")

Private Members

cl::Context context_

cl::CommandQueue commandQueue_

size_t maxWorkgroupSize_

size_t maxNumWorkitems_

size_t entriesPerStream_

I3CLSimQueue<bunch_t> stepQueue_

boost::thread harvesterThread_

bool run_

double domArea_

double stepLength_

std::pair<double, double> minimumRefractiveIndex_

group, phase

clsim::tabulator::AxesConstPtr axes_

std::vector<float> binContent_

std::vector<float> squaredWeights_

uint64_t numPhotons_

double sumOfPhotonWeights_

double spectralBiasFactor_

number of Photonics photons represented by each clsim photon

class I3CLSimTabulatorModule : public I3Module

Public Functions

I3CLSimTabulatorModule(const I3Context&)

virtual ~I3CLSimTabulatorModule()

void Configure()

void DAQ(I3FramePtr)

void Finish()

Public Members

boost::mutex mutex

boost::condition_variable cv

Private Functions

void HarvestSteps()

Harvest steps and feed them to the tabulator.

SET_LOGGER ("I3CLSimTabulatorModule")

Private Members

std::string mctreeName_

std::string flasherPulseSeriesName_

I3LightSourceParameterizationSeries parameterizationList_

I3RandomServicePtr randomService_

I3CLSimFunctionConstPtr wavelengthGenerationBias_

I3CLSimMediumPropertiesConstPtr mediumProperties_

I3CLSimFunctionConstPtr angularAcceptance_

I3CLSimSpectrumTableConstPtr spectrumTable_

I3CLSimOpenCLDeviceSeries openCLDeviceList_

double referenceArea_

size_t photonsPerBunch_

size_t entriesPerPhoton_

bool recordErrors_

I3LightSourceToStepConverterAsyncPtr particleToStepsConverter_

boost::scoped_ptr<I3CLSimStepToTableConverter> tabulator_

std::string tablePath_

boost::python::dict tableHeader_

clsim::tabulator::AxesPtr axes_

boost::thread stepHarvester_

struct I3CLSimTabulatorModule::[anonymous] semaphore

I3CLSimQueue<I3ParticleConstPtr> sourceQueue_

bool run_

struct I3MCHitConverterWithIDs

#include <I3MCHitConverterWithIDs.h>

The default tableio I3MCHit converter does not store te hit id. This is a similar converter that also stores hit ids and particle ids.

Public Types

typedef I3MCHit booked_type

typedef booked_type value_type

Public Functions

void AddFields(I3TableRowDescriptionPtr desc, const booked_type& = booked_type())

void FillSingleRow(const booked_type &dl, I3TableRowPtr row)

class I3MuonSlicer : public I3ConditionalModule

#include <I3MuonSlicer.h>

takes an MC tree with a muon that has been propagated using MMC (i.e. a muon with cascades as daughter particles). It chops the muon track into slices with estimated energies. All the original muon will be retained and all slices will be added to the muon as daughter particles. So you will end up with a long muon track with cascades and shorter muons as daughter particles. Make sure your light propagation knows how to handle this (you run the risk of having light generated twice, once for the parent (=long) track and again for the daughters muon tracks).

I3CLSimModule knows how to deal with these tracks.

The muon starting/stopping points and energies are taken from an I3MMCTrackList object in case it exists and a name is configured.

Public Functions

I3MuonSlicer(const I3Context &ctx)

Builds an instance of this class

virtual ~I3MuonSlicer()

Destroys an instance of this class

virtual void Configure()

This module takes a configuration parameter and so it must be configured.

virtual void DAQ(I3FramePtr frame)

The module needs to process Physics frames

Private Functions

I3MuonSlicer()

I3MuonSlicer(const I3MuonSlicer&)

I3MuonSlicer &operator=(const I3MuonSlicer&)

SET_LOGGER ("I3MuonSlicer")

Private Members

std::string inputMCTreeName_

Parameter: Name of the I3MCTree frame object.

std::string MMCTrackListName_

Parameter: Name of the I3MMCTrackList frame object.

std::string outputMCTreeName_

Parameter: Name of the output I3MCTree frame object. If identical to the input or empty, the input object will be replaced.

class I3MuonSliceRemoverAndPulseRelabeler : public I3ConditionalModule

Removes muon slices in I3MCTree objects as written by I3MuonSlicer and re-labels I3MCPEs generated from these so that they refer back to the original muon.

Applying this after having generated I3MCPEs using clsim will generate an MCTree fully compatible with what ppc would generate (i.e. the light will look light it had been created by the original muon)

TODO: THIS IS NOT READY FOR USE YET!

Public Functions

I3MuonSliceRemoverAndPulseRelabeler(const I3Context &ctx)

Builds an instance of this class

virtual ~I3MuonSliceRemoverAndPulseRelabeler()

Destroys an instance of this class

virtual void Configure()

This module takes a configuration parameter and so it must be configured.

virtual void DAQ(I3FramePtr frame)

The module needs to process Physics frames

Private Functions

I3MuonSliceRemoverAndPulseRelabeler()

I3MuonSliceRemoverAndPulseRelabeler(const I3MuonSliceRemoverAndPulseRelabeler&)

I3MuonSliceRemoverAndPulseRelabeler &operator=(const I3MuonSliceRemoverAndPulseRelabeler&)

SET_LOGGER ("I3MuonSliceRemoverAndPulseRelabeler")

Private Members

std::string inputMCTreeName_

Parameter: Name of the input I3MCTree frame object as written by I3MuonSlicer.

std::string oldMCTreeName_

Parameter: Name of the original MCTree. Will be used for sanity checks.

std::string inputMCPESeriesMapName_

Parameter: Name of the input I3MCPESeriesMap frame object.

std::string outputMCPESeriesMapName_

Parameter: Name of the output I3MCPESeriesMap frame object (can be identical to the input).

std::string inputPhotonSeriesMapName_

Parameter: Name of the input I3PhotonSeriesMap frame object.

std::string outputPhotonSeriesMapName_

Parameter: Name of the output I3PhotonSeriesMap frame object (can be identical to the input).

class I3PhotonToMCHitConverterForMultiPMT : public I3ConditionalModule

#include <I3PhotonToMCHitConverterForMultiPMT.h>

This module uses PMT and OM acceptance information from the multiPMT-patched I3Geometry class to convert from an I3PhotonMap to an I3MCHitSeriesMapMap (or I3MCHitSeriesMultiOMMap) which can be passed to a PMT and readout simulation.

Functionality from this module should probably integrated in the I3PhotonToMCHitConverter module..

Public Functions

I3PhotonToMCHitConverterForMultiPMT(const I3Context &ctx)

Constructor.

virtual ~I3PhotonToMCHitConverterForMultiPMT()

Destructor.

void Configure()

Configure this module.

void DAQ(I3FramePtr frame)

We subscribe to physics events.

Parameters:

frame – The frame

void Finish()

To clean up.

Private Functions

SET_LOGGER ("I3PhotonToMCHitConverterForMultiPMT")

The logger can also be used for this module.

Private Members

I3RandomServicePtr randomService_

Parameter:A random number generating service (derived from I3RandomService).

std::string inputPhotonSeriesMapName_

Parameter: Name of the input I3PhotonSeriesMap frame object

std::string outputMCHitSeriesMapName_

Parameter: Name of the I3MCHitSeriesMap which will be put into the frame.

std::string MCTreeName_

Parameter: Name of the I3MCTree frame object. All photon particle IDs are checked against this tree.

class I3ShadowedPhotonRemoverModule : public I3ConditionalModule

#include <I3ShadowedPhotonRemoverModule.h>

Public Functions

I3ShadowedPhotonRemoverModule(const I3Context &ctx)

~I3ShadowedPhotonRemoverModule()

virtual void Configure()

void DAQ(I3FramePtr frame)

Private Functions

I3ShadowedPhotonRemoverModule()

I3ShadowedPhotonRemoverModule(const I3ShadowedPhotonRemoverModule&)

I3ShadowedPhotonRemoverModule &operator=(const I3ShadowedPhotonRemoverModule&)

SET_LOGGER ("I3ShadowedPhotonRemoverModule")

Private Members

std::string inputPhotonSeriesMapName_

Parameter: Name of the input I3PhotonSeriesMap frame object.

std::string outputPhotonSeriesMapName_

std::string cylinder_map_name_

double distance_

I3CylinderMapConstPtr cylinder_map_

class I3TauSanitizer : public I3ConditionalModule

Sets taus with length==0 or NaN to shape “Dark”.

Public Functions

I3TauSanitizer(const I3Context &ctx)

Builds an instance of this class

virtual ~I3TauSanitizer()

Destroys an instance of this class

virtual void Configure()

This module takes a configuration parameter and so it must be configured.

virtual void DAQ(I3FramePtr frame)

The module needs to process Physics frames

Private Functions

I3TauSanitizer()

I3TauSanitizer(const I3TauSanitizer&)

I3TauSanitizer &operator=(const I3TauSanitizer&)

SET_LOGGER ("I3TauSanitizer")

Private Members

std::string inputMCTreeName_

Parameter: Name of the I3MCTree frame object.

std::string outputMCTreeName_

Parameter: Name of the output I3MCTree frame object. If identical to the input or empty, the input object will be replaced.

class Image : public cl::Memory

#include <opencl.hpp>

C++ base class for Image Memory objects.

See Memory for details about copy semantics, etc.

See also

Memory

Subclassed by cl::Image1D, cl::Image1DArray, cl::Image1DBuffer, cl::Image2D, cl::Image2DArray, cl::Image3D, cl::ImageGL

Public Functions

template<typename T>
inline cl_int getImageInfo(cl_image_info name, T *param) const

Wrapper for clGetImageInfo().

template<cl_image_info name>
inline detail::param_traits<detail::cl_image_info, name>::param_type getImageInfo(cl_int *err = NULL) const

Wrapper for clGetImageInfo() that returns by value.

Protected Functions

inline Image()

Default constructor - initializes to NULL.

inline explicit Image(const cl_mem &image, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Image(const Image &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image &operator=(const Image &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image (Image &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image &operator=(Image &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image1D : public cl::Image

#include <opencl.hpp>

Class interface for 1D Image Memory objects.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline Image1D(const Context &context, cl_mem_flags flags, ImageFormat format, size_type width, void *host_ptr = NULL, cl_int *err = NULL)

Constructs a 1D Image in a specified context.

Wraps clCreateImage().

inline Image1D()

Default constructor - initializes to NULL.

inline explicit Image1D(const cl_mem &image1D, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image1D &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Image1D(const Image1D &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image1D &operator=(const Image1D &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image1D (Image1D &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image1D &operator=(Image1D &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image1DArray : public cl::Image

#include <opencl.hpp>

Image interface for arrays of 1D images.

Public Functions

inline Image1DArray(const Context &context, cl_mem_flags flags, ImageFormat format, size_type arraySize, size_type width, size_type rowPitch, void *host_ptr = NULL, cl_int *err = NULL)

inline Image1DArray()

inline explicit Image1DArray(const cl_mem &imageArray, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image1DArray &operator=(const cl_mem &rhs)

inline Image1DArray(const Image1DArray &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image1DArray &operator=(const Image1DArray &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image1DArray (Image1DArray &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image1DArray &operator=(Image1DArray &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image1DBuffer : public cl::Image

#include <opencl.hpp>

Image interface for 1D buffer images.

Public Functions

inline Image1DBuffer(const Context &context, cl_mem_flags flags, ImageFormat format, size_type width, const Buffer &buffer, cl_int *err = NULL)

inline Image1DBuffer()

inline explicit Image1DBuffer(const cl_mem &image1D, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image1DBuffer &operator=(const cl_mem &rhs)

inline Image1DBuffer(const Image1DBuffer &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image1DBuffer &operator=(const Image1DBuffer &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image1DBuffer (Image1DBuffer &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image1DBuffer &operator=(Image1DBuffer &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image2D : public cl::Image

#include <opencl.hpp>

Class interface for 2D Image Memory objects.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline Image2D(const Context &context, cl_mem_flags flags, ImageFormat format, size_type width, size_type height, size_type row_pitch = 0, void *host_ptr = NULL, cl_int *err = NULL)

Constructs a 2D Image in a specified context.

Wraps clCreateImage().

inline Image2D(const Context &context, ImageFormat format, const Buffer &sourceBuffer, size_type width, size_type height, size_type row_pitch = 0, cl_int *err = nullptr)

Constructs a 2D Image from a buffer.

Wraps clCreateImage().

Note

This will share storage with the underlying buffer.

inline Image2D(const Context &context, cl_channel_order order, const Image &sourceImage, cl_int *err = nullptr)

Constructs a 2D Image from an image.

The image will be created matching with a descriptor matching the source.

Wraps clCreateImage().

Note

This will share storage with the underlying image but may reinterpret the channel order and type.

Parameters:

order – is the channel order to reinterpret the image data as. The channel order may differ as described in the OpenCL 2.0 API specification.

inline Image2D()

Default constructor - initializes to NULL.

inline explicit Image2D(const cl_mem &image2D, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image2D &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Image2D(const Image2D &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image2D &operator=(const Image2D &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image2D (Image2D &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image2D &operator=(Image2D &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image2DArray : public cl::Image

#include <opencl.hpp>

Image interface for arrays of 2D images.

Public Functions

inline Image2DArray(const Context &context, cl_mem_flags flags, ImageFormat format, size_type arraySize, size_type width, size_type height, size_type rowPitch, size_type slicePitch, void *host_ptr = NULL, cl_int *err = NULL)

inline Image2DArray()

inline explicit Image2DArray(const cl_mem &imageArray, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image2DArray &operator=(const cl_mem &rhs)

inline Image2DArray(const Image2DArray &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image2DArray &operator=(const Image2DArray &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image2DArray (Image2DArray &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image2DArray &operator=(Image2DArray &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Image3D : public cl::Image

#include <opencl.hpp>

Class interface for 3D Image Memory objects.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline Image3D(const Context &context, cl_mem_flags flags, ImageFormat format, size_type width, size_type height, size_type depth, size_type row_pitch = 0, size_type slice_pitch = 0, void *host_ptr = NULL, cl_int *err = NULL)

Constructs a 3D Image in a specified context.

Wraps clCreateImage().

inline Image3D()

Default constructor - initializes to NULL.

inline explicit Image3D(const cl_mem &image3D, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline Image3D &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Image3D(const Image3D &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Image3D &operator=(const Image3D &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Image3D (Image3D &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Image3D &operator=(Image3D &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

struct ImageFormat : public cl_image_format

#include <opencl.hpp>

Adds constructors and member functions for cl_image_format.

ImageFormat

See also

cl_image_format

Public Functions

inline ImageFormat()

Default constructor - performs no initialization.

inline ImageFormat(cl_channel_order order, cl_channel_type type)

Initializing constructor.

inline ImageFormat(const ImageFormat &other)

Copy constructor.

inline ImageFormat &operator=(const ImageFormat &rhs)

Assignment operator.

class ImageGL : public cl::Image

#include <opencl.hpp>

general image interface for GL interop. We abstract the 2D and 3D GL images into a single instance here that wraps all GL sourced images on the grounds that setup information was performed by OpenCL anyway.

Public Functions

inline ImageGL(const Context &context, cl_mem_flags flags, cl_GLenum target, cl_GLint miplevel, cl_GLuint texobj, cl_int *err = NULL)

inline ImageGL()

inline explicit ImageGL(const cl_mem &image, bool retainObject = false)

Constructor from cl_mem - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. See Memory for further details.

inline ImageGL &operator=(const cl_mem &rhs)

inline ImageGL(const ImageGL &img)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline ImageGL &operator=(const ImageGL &img)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  ImageGL (ImageGL &&img) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline ImageGL &operator=(ImageGL &&img)

Move assignment to forward move to the superclass correctly. Required for MSVC.

class Kernel : public cl::detail::Wrapper<cl_kernel>

#include <opencl.hpp>

Class interface for cl_kernel.

See also

cl_kernel

Note

Copies of these objects are shallow, meaning that the copy will refer to the same underlying cl_kernel as the original. For details, see clRetainKernel() and clReleaseKernel().

Public Functions

inline Kernel(const Program &program, const char *name, cl_int *err = NULL)

inline Kernel()

Default constructor - initializes to NULL.

inline explicit Kernel(const cl_kernel &kernel, bool retainObject = false)

Constructor from cl_kernel - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. This effectively transfers ownership of a refcount on the cl_kernel into the new Kernel object.

inline Kernel &operator=(const cl_kernel &rhs)

Assignment operator from cl_kernel - takes ownership.

This effectively transfers ownership of a refcount on the rhs and calls clReleaseKernel() on the value previously held by this instance.

inline Kernel(const Kernel &kernel)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Kernel &operator=(const Kernel &kernel)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Kernel (Kernel &&kernel) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Kernel &operator=(Kernel &&kernel)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_kernel_info name, T *param) const

template<cl_kernel_info name>
inline detail::param_traits<detail::cl_kernel_info, name>::param_type getInfo(cl_int *err = NULL) const

template<typename T>
inline cl_int getArgInfo(cl_uint argIndex, cl_kernel_arg_info name, T *param) const

template<cl_kernel_arg_info name>
inline detail::param_traits<detail::cl_kernel_arg_info, name>::param_type getArgInfo(cl_uint argIndex, cl_int *err = NULL) const

template<typename T>
inline cl_int getWorkGroupInfo(const Device &device, cl_kernel_work_group_info name, T *param) const

template<cl_kernel_work_group_info name>
inline detail::param_traits<detail::cl_kernel_work_group_info, name>::param_type getWorkGroupInfo(const Device &device, cl_int *err = NULL) const

inline cl_int getSubGroupInfo(const cl::Device &dev, cl_kernel_sub_group_info name, const cl::NDRange &range, size_type *param) const

template<cl_kernel_sub_group_info name>
inline size_type getSubGroupInfo(const cl::Device &dev, const cl::NDRange &range, cl_int *err = NULL) const

template<typename T, class D>
inline cl_int setArg(cl_uint index, const cl::pointer<T, D> &argPtr)

setArg overload taking a shared_ptr type

template<typename T, class Alloc>
inline cl_int setArg(cl_uint index, const cl::vector<T, Alloc> &argPtr)

setArg overload taking a vector type.

template<typename T>
inline std::enable_if<std::is_pointer<T>::value, cl_int>::type setArg(cl_uint index, const T argPtr)

setArg overload taking a pointer type

template<typename T>
inline std::enable_if<!std::is_pointer<T>::value, cl_int>::type setArg(cl_uint index, const T &value)

setArg overload taking a POD type

inline cl_int setArg(cl_uint index, size_type size, const void *argPtr)

inline cl_int setSVMPointers(const vector<void*> &pointerList)

Specify a vector of SVM pointers that the kernel may access in addition to its arguments.

template<int ArrayLength>
inline cl_int setSVMPointers(const std::array<void*, ArrayLength> &pointerList)

Specify a std::array of SVM pointers that the kernel may access in addition to its arguments.

inline cl_int enableFineGrainedSystemSVM(bool svmEnabled)

Enable fine-grained system SVM.

See also

clSetKernelExecInfo

Note

It is only possible to enable fine-grained system SVM if all devices in the context associated with kernel support it.

Parameters:

svmEnabled – True if fine-grained system SVM is requested. False otherwise.

Returns:

CL_SUCCESS if the function was executed succesfully. CL_INVALID_OPERATION if no devices in the context support fine-grained system SVM.

template<int index, int ArrayLength, class D, typename T0, typename T1, typename ...Ts>
inline void setSVMPointersHelper(std::array<void*, ArrayLength> &pointerList, const pointer<T0, D> &t0, const pointer<T1, D> &t1, Ts&... ts)

template<int index, int ArrayLength, typename T0, typename T1, typename ...Ts>
inline std::enable_if<std::is_pointer<T0>::value, void>::type setSVMPointersHelper(std::array<void*, ArrayLength> &pointerList, T0 t0, T1 t1, Ts... ts)

template<int index, int ArrayLength, typename T0, class D>
inline void setSVMPointersHelper(std::array<void*, ArrayLength> &pointerList, const pointer<T0, D> &t0)

template<int index, int ArrayLength, typename T0>
inline std::enable_if<std::is_pointer<T0>::value, void>::type setSVMPointersHelper(std::array<void*, ArrayLength> &pointerList, T0 t0)

template<typename T0, typename ...Ts>
inline cl_int setSVMPointers(const T0 &t0, Ts&... ts)

template<typename T>
inline cl_int setExecInfo(cl_kernel_exec_info param_name, const T &val)

template<cl_kernel_exec_info name>
inline cl_int setExecInfo(typename detail::param_traits<detail::cl_kernel_exec_info, name>::param_type &val)

inline Kernel clone()

Make a deep copy of the kernel object including its arguments.

Returns:

A new kernel object with internal state entirely separate from that of the original but with any arguments set on the original intact.

template<typename T, class Enable = void>
struct KernelArgumentHandler

#include <opencl.hpp>

template<>
struct KernelArgumentHandler<cl::DeviceCommandQueue, void>

#include <opencl.hpp>

Public Static Functions

static inline size_type size(const cl::DeviceCommandQueue&)

static inline const cl_command_queue *ptr(const cl::DeviceCommandQueue &value)

template<>
struct KernelArgumentHandler<LocalSpaceArg, void>

#include <opencl.hpp>

Public Static Functions

static inline size_type size(const LocalSpaceArg &value)

static inline const void *ptr(const LocalSpaceArg&)

template<typename T>
struct KernelArgumentHandler<T, typename std::enable_if<std::is_base_of<cl::Memory, T>::value>::type>

#include <opencl.hpp>

Public Static Functions

static inline size_type size(const T&)

static inline const cl_mem *ptr(const T &value)

template<typename T>
struct KernelArgumentHandler<T, typename std::enable_if<!std::is_base_of<cl::Memory, T>::value>::type>

#include <opencl.hpp>

Public Static Functions

static inline size_type size(const T&)

static inline const T *ptr(const T &value)

template<typename ...Ts>
class KernelFunctor

#include <opencl.hpp>

Type safe kernel functor.

Public Types

typedef Event result_type

Return type of the functor.

Public Functions

inline KernelFunctor(Kernel kernel)

inline KernelFunctor(const Program &program, const string name, cl_int *err = NULL)

inline Event operator()(const EnqueueArgs &args, Ts... ts)

Enqueue kernel.

Parameters:

• args – Launch parameters of the kernel.

• t0... – List of kernel arguments based on the template type of the functor.

inline Event operator()(const EnqueueArgs &args, Ts... ts, cl_int &error)

Enqueue kernel with support for error code.

Parameters:

• args – Launch parameters of the kernel.

• t0... – List of kernel arguments based on the template type of the functor.

• error – Out parameter returning the error code from the execution.

inline cl_int setSVMPointers(const vector<void*> &pointerList)

template<typename T0, typename ...T1s>
inline cl_int setSVMPointers(const T0 &t0, T1s&... ts)

inline Kernel getKernel()

Private Functions

template<int index, typename T0, typename ...T1s>
inline void setArgs(T0 &&t0, T1s&&... t1s)

template<int index, typename T0>
inline void setArgs(T0 &&t0)

template<int index>
inline void setArgs()

Private Members

Kernel kernel_

struct LightSourceData_t

Public Members

bool isBarrier

I3FlasherPulse flasherPulse

uint32_t identifier

uint64_t numPhotonsWithBias

class LinearAxis : public clsim::tabulator::Axis

#include <Axis.h>

Public Functions

LinearAxis(double min, double max, unsigned n_bins)

virtual ~LinearAxis()

virtual double Transform(double value) const

Transform value from linear into nonlinear space.

virtual double InverseTransform(double value) const

Transform value from nonlinear into linear space.

virtual std::string GetTransformCode(const std::string &varName) const

Generate OpenCL code to transform variable from linear into nonlinear space.

virtual std::string GetInverseTransformCode(const std::string &varName) const

Generate OpenCL code to transform variable from nonlinear into linear space.

struct LocalSpaceArg

#include <opencl.hpp>

Local address wrapper for use with Kernel::setArg.

Public Members

size_type size_

template<typename ...Ts>
struct make_kernel

#include <opencl.hpp>

Backward compatibility class to ensure that cl.hpp code works with opencl.hpp. Please use KernelFunctor directly.

Public Types

typedef KernelFunctor<Ts...> FunctorType

typedef Event result_type

Return type of the functor.

typedef Event type_(const EnqueueArgs&, Ts...)

Function signature of kernel functor with no event dependency.

Public Functions

inline make_kernel(const Program &program, const string name, cl_int *err = NULL)

inline make_kernel(const Kernel kernel)

inline Event operator()(const EnqueueArgs &enqueueArgs, Ts... args)

Public Members

FunctorType functor_

class MakeSteps_visitor : public boost::static_visitor<std::pair<I3SimStepSeriesConstPtr, bool>>

Public Functions

MakeSteps_visitor(uint64_t &rngState, uint32_t rngA, uint64_t maxNumStepsPerStepSeries, GenerateStepPreCalculator &preCalc)

template<typename T>
std::pair<I3SimStepSeriesConstPtr, bool> operator()(T &data) const

template<>
std::pair<I3SimStepSeriesConstPtr, bool> operator()(I3CLSimLightSourceToStepConverterPPC::BarrierData_t &data) const

Private Functions

void FillStep(I3CLSimLightSourceToStepConverterPPC::CascadeStepData_t &data, I3SimStep &newStep, uint64_t photonsPerStep, double particleDir_x, double particleDir_y, double particleDir_z) const

void FillStep(I3CLSimLightSourceToStepConverterPPC::MuonStepData_t &data, I3SimStep &newStep, uint64_t photonsPerStep, double particleDir_x, double particleDir_y, double particleDir_z) const

Private Members

uint64_t &rngState_

uint32_t rngA_

uint64_t maxNumStepsPerStepSeries_

GenerateStepPreCalculator &preCalc_

class Memory : public cl::detail::Wrapper<cl_mem>

#include <opencl.hpp>

Class interface for cl_mem.

See also

cl_mem

Note

Copies of these objects are shallow, meaning that the copy will refer to the same underlying cl_mem as the original. For details, see clRetainMemObject() and clReleaseMemObject().

Subclassed by cl::Buffer, cl::Image, cl::Pipe

Public Functions

inline Memory()

Default constructor - initializes to NULL.

inline explicit Memory(const cl_mem &memory, bool retainObject)

Constructor from cl_mem - takes ownership.

Optionally transfer ownership of a refcount on the cl_mem into the new Memory object.

See Memory for further details.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline Memory &operator=(const cl_mem &rhs)

Assignment operator from cl_mem - takes ownership.

This effectively transfers ownership of a refcount on the rhs and calls clReleaseMemObject() on the value previously held by this instance.

inline Memory(const Memory &mem)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Memory &operator=(const Memory &mem)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Memory (Memory &&mem) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Memory &operator=(Memory &&mem)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_mem_info name, T *param) const

Wrapper for clGetMemObjectInfo().

template<cl_mem_info name>
inline detail::param_traits<detail::cl_mem_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetMemObjectInfo() that returns by value.

inline cl_int setDestructorCallback (void(CL_CALLBACK *pfn_notify)(cl_mem, void *), void *user_data=NULL)

Registers a callback function to be called when the memory object is no longer needed.

Wraps clSetMemObjectDestructorCallback().

Repeated calls to this function, for a given cl_mem value, will append to the list of functions called (in reverse order) when memory object’s resources are freed and the memory object is deleted.

Note

The registered callbacks are associated with the underlying cl_mem value - not the Memory class instance.

class metric

Public Functions

inline metric()

inline void update(double value)

inline size_t count() const

inline double mean() const

inline double variance() const

Private Members

size_t count_

double mean_

double mean2_

class metric

Public Functions

inline metric()

inline void update(double value)

inline size_t count() const

inline double mean() const

inline double variance() const

Private Members

size_t count_

double mean_

double mean2_

template<typename CBRNG>
class MicroURNG

#include <MicroURNG.hpp>

Given a CBRNG whose ctr_type has an unsigned integral value_type, MicroURNG<CBRNG>(c, k) is a type that satisfies the requirements of a C++0x Uniform Random Number Generator.

The intended purpose is for a MicroURNG to be passed as an argument to a C++0x Distribution, e.g., std::normal_distribution. See examples/MicroURNG.cpp.

The MicroURNG functor has a period of “only”

ctr_type.size()*2^32,

after which it will silently repeat.

The high 32 bits of the highest word in the counter c, passed to the constructor must be zero. MicroURNG uses these bits to “count”.

Older versions of the library permitted a second template parameter by which the caller could control the number of bits devoted to the URNG’s internal counter. This flexibility has been disabled because URNGs created with different numbers of counter bits could, conceivably “collide”.

typedef ?someCBRNG? RNG;
RNG::ctr_type c = ...; // under application control
RNG::key_type k = ...; // 
std::normal_distribution<float> nd;
MicroURNG<RNG> urng(c, k);
for(???){
  ...
  nd(urng);  // may be called several hundred times with BITS=10
  ...
}

Public Types

typedef CBRNG cbrng_type

typedef cbrng_type::ctr_type ctr_type

typedef cbrng_type::key_type key_type

typedef cbrng_type::ukey_type ukey_type

typedef ctr_type::value_type result_type

Public Functions

R123_STATIC_ASSERT (std::numeric_limits< result_type >::digits >=BITS,"The result_type must have at least 32 bits")

inline result_type operator()()

inline MicroURNG(cbrng_type _b, ctr_type _c0, ukey_type _uk)

inline MicroURNG(ctr_type _c0, ukey_type _uk)

inline const ctr_type &counter() const

inline void reset(ctr_type _c0, ukey_type _uk)

Public Static Functions

static inline R123_CONSTEXPR result_type min R123_NO_MACRO_SUBST ()

static inline R123_CONSTEXPR result_type max R123_NO_MACRO_SUBST ()

Public Static Attributes

static const int BITS = 32

static const result_type _Min = 0

static const result_type _Max = ~((result_type)0)

Private Functions

inline void chkhighbits()

Private Members

cbrng_type b

ctr_type c0

key_type k

R123_ULONG_LONG n

size_t last_elem

ctr_type rdata

struct MuonStepData_t

Public Members

I3Particle particle

uint32_t particleIdentifier

uint64_t photonsPerStep

uint64_t numSteps

uint64_t numPhotonsInLastStep

bool stepIsCascadeLike

double length

class NDRange

#include <opencl.hpp>

Class interface for specifying NDRange values.

Public Functions

inline NDRange()

Default constructor - resulting range has zero dimensions.

inline NDRange(size_type size0)

Constructs one-dimensional range.

inline NDRange(size_type size0, size_type size1)

Constructs two-dimensional range.

inline NDRange(size_type size0, size_type size1, size_type size2)

Constructs three-dimensional range.

inline operator const size_type*() const

Conversion operator to const size_type *.

Returns:

a pointer to the size of the first dimension.

inline size_type dimensions() const

Queries the number of dimensions in the range.

inline size_type size() const

Returns the size of the object in bytes based on the.

inline size_type *get()

inline const size_type *get() const

Private Members

size_type sizes_[3]

cl_uint dimensions_

template<typename enum_type, cl_int Name>
struct param_traits

#include <opencl.hpp>

struct particleCacheEntry

#include <I3CLSimModule.h>

Public Members

std::size_t frameListEntry

uint64_t particleMajorID

int particleMinorID

double timeShift

class Pipe : public cl::Memory

#include <opencl.hpp>

Class interface for Pipe Memory Objects.

See Memory for details about copy semantics, etc.

See also

Memory

Public Functions

inline Pipe(const Context &context, cl_uint packet_size, cl_uint max_packets, cl_int *err = NULL)

Constructs a Pipe in a specified context.

Wraps clCreatePipe().

Parameters:

• context – Context in which to create the pipe.

• flags – Bitfield. Only CL_MEM_READ_WRITE and CL_MEM_HOST_NO_ACCESS are valid.

• packet_size – Size in bytes of a single packet of the pipe.

• max_packets – Number of packets that may be stored in the pipe.

inline Pipe(cl_uint packet_size, cl_uint max_packets, cl_int *err = NULL)

Constructs a Pipe in a the default context.

Wraps clCreatePipe().

Parameters:

• flags – Bitfield. Only CL_MEM_READ_WRITE and CL_MEM_HOST_NO_ACCESS are valid.

• packet_size – Size in bytes of a single packet of the pipe.

• max_packets – Number of packets that may be stored in the pipe.

inline Pipe()

Default constructor - initializes to NULL.

inline explicit Pipe(const cl_mem &pipe, bool retainObject = false)

Constructor from cl_mem - takes ownership.

See Memory for further details.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline Pipe &operator=(const cl_mem &rhs)

Assignment from cl_mem - performs shallow copy.

See Memory for further details.

inline Pipe(const Pipe &pipe)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Pipe &operator=(const Pipe &pipe)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Pipe (Pipe &&pipe) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Pipe &operator=(Pipe &&pipe)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_pipe_info name, T *param) const

Wrapper for clGetMemObjectInfo().

template<cl_pipe_info name>
inline detail::param_traits<detail::cl_pipe_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetMemObjectInfo() that returns by value.

class Platform : public cl::detail::Wrapper<cl_platform_id>

#include <opencl.hpp>

Class interface for cl_platform_id.

See also

cl_platform_id

Note

Copies of these objects are inexpensive, since they don’t ‘own’ any underlying resources or data structures.

Public Functions

inline Platform()

Default constructor - initializes to NULL.

inline explicit Platform(const cl_platform_id &platform, bool retainObject = false)

Constructor from cl_platform_id.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. This simply copies the platform ID value, which is an inexpensive operation.

inline Platform &operator=(const cl_platform_id &rhs)

Assignment operator from cl_platform_id.

This simply copies the platform ID value, which is an inexpensive operation.

template<typename T>
inline cl_int getInfo(cl_platform_info name, T *param) const

Wrapper for clGetPlatformInfo().

template<cl_platform_info name>
inline detail::param_traits<detail::cl_platform_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetPlatformInfo() that returns by value.

inline cl_int getDevices(cl_device_type type, vector<Device> *devices) const

Gets a list of devices for this platform.

Wraps clGetDeviceIDs().

inline cl_int unloadCompiler()

Wrapper for clUnloadCompiler().

Public Static Functions

static inline Platform getDefault(cl_int *errResult = NULL)

static inline Platform setDefault(const Platform &default_platform)

Modify the default platform to be used by subsequent operations. Will only set the default if no default was previously created.

Returns:

updated default platform. Should be compared to the passed value to ensure that it was updated.

static inline cl_int get(vector<Platform> *platforms)

Gets a list of available platforms.

Wraps clGetPlatformIDs().

static inline cl_int get(Platform *platform)

Gets the first available platform.

Wraps clGetPlatformIDs(), returning the first result.

static inline Platform get(cl_int *errResult = NULL)

Gets the first available platform, returning it by value.

Returns:

Returns a valid platform if one is available. If no platform is available will return a null platform. Throws an exception if no platforms are available or an error condition occurs. Wraps clGetPlatformIDs(), returning the first result.

Private Static Functions

static inline void makeDefault()

Create the default context.

This sets default_ and default_error_. It does not throw cl::Error.

static inline void makeDefaultProvided(const Platform &p)

Create the default platform from a provided platform.

This sets default_. It does not throw cl::Error.

Private Static Attributes

static std::once_flag default_initialized_

static Platform default_

static cl_int default_error_ = CL_SUCCESS

class PowerAxis : public clsim::tabulator::Axis

#include <Axis.h>

Public Functions

PowerAxis(double min, double max, unsigned n_bins, unsigned power = 1)

virtual ~PowerAxis()

inline unsigned GetPower() const

virtual double Transform(double value) const

Transform value from linear into nonlinear space.

virtual double InverseTransform(double value) const

Transform value from nonlinear into linear space.

virtual std::string GetTransformCode(const std::string &varName) const

Generate OpenCL code to transform variable from linear into nonlinear space.

virtual std::string GetInverseTransformCode(const std::string &varName) const

Generate OpenCL code to transform variable from nonlinear into linear space.

Private Members

unsigned power_

class Program : public cl::detail::Wrapper<cl_program>

#include <opencl.hpp>

Program interface that implements cl_program.

Public Types

typedef vector<vector<unsigned char>> Binaries

typedef vector<string> Sources

Public Functions

inline Program(const string &source, bool build = false, cl_int *err = NULL)

inline Program(const Context &context, const string &source, bool build = false, cl_int *err = NULL)

inline Program(const Sources &sources, cl_int *err = NULL)

Create a program from a vector of source strings and the default context. Does not compile or link the program.

inline Program(const Context &context, const Sources &sources, cl_int *err = NULL)

Create a program from a vector of source strings and a provided context. Does not compile or link the program.

inline Program(const vector<char> &IL, bool build = false, cl_int *err = NULL)

Program constructor to allow construction of program from SPIR-V or another IL. Valid for either OpenCL >= 2.1 or when CL_HPP_USE_IL_KHR is defined.

inline Program(const Context &context, const vector<char> &IL, bool build = false, cl_int *err = NULL)

Program constructor to allow construction of program from SPIR-V or another IL for a specific context. Valid for either OpenCL >= 2.1 or when CL_HPP_USE_IL_KHR is defined.

inline Program(const Context &context, const vector<Device> &devices, const Binaries &binaries, vector<cl_int> *binaryStatus = NULL, cl_int *err = NULL)

Construct a program object from a list of devices and a per-device list of binaries.

Parameters:

• context – A valid OpenCL context in which to construct the program.

• devices – A vector of OpenCL device objects for which the program will be created.

• binaries – A vector of pairs of a pointer to a binary object and its length.

• binaryStatus – An optional vector that on completion will be resized to match the size of binaries and filled with values to specify if each binary was successfully loaded. Set to CL_SUCCESS if the binary was successfully loaded. Set to CL_INVALID_VALUE if the length is 0 or the binary pointer is NULL. Set to CL_INVALID_BINARY if the binary provided is not valid for the matching device.

• err – if non-NULL will be set to CL_SUCCESS on successful operation or one of the following errors: CL_INVALID_CONTEXT if context is not a valid context. CL_INVALID_VALUE if the length of devices is zero; or if the length of binaries does not match the length of devices; or if any entry in binaries is NULL or has length 0. CL_INVALID_DEVICE if OpenCL devices listed in devices are not in the list of devices associated with context. CL_INVALID_BINARY if an invalid program binary was encountered for any device. binaryStatus will return specific status for each device. CL_OUT_OF_HOST_MEMORY if there is a failure to allocate resources required by the OpenCL implementation on the host.

inline Program(const Context &context, const vector<Device> &devices, const string &kernelNames, cl_int *err = NULL)

Create program using builtin kernels.

Parameters:

kernelNames – Semi-colon separated list of builtin kernel names

inline Program()

inline explicit Program(const cl_program &program, bool retainObject = false)

Constructor from cl_program - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions.

inline Program &operator=(const cl_program &rhs)

inline Program(const Program &program)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Program &operator=(const Program &program)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Program (Program &&program) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Program &operator=(Program &&program)

Move assignment to forward move to the superclass correctly. Required for MSVC.

inline cl_int build (const vector< Device > &devices, const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL) const

inline cl_int build (const Device &device, const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL) const

inline cl_int build (const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL) const

inline cl_int compile (const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL) const

template<typename T>
inline cl_int getInfo(cl_program_info name, T *param) const

template<cl_program_info name>
inline detail::param_traits<detail::cl_program_info, name>::param_type getInfo(cl_int *err = NULL) const

template<typename T>
inline cl_int getBuildInfo(const Device &device, cl_program_build_info name, T *param) const

template<cl_program_build_info name>
inline detail::param_traits<detail::cl_program_build_info, name>::param_type getBuildInfo(const Device &device, cl_int *err = NULL) const

template<cl_program_build_info name>
inline vector<std::pair<cl::Device, typename detail::param_traits<detail::cl_program_build_info, name>::param_type>> getBuildInfo(cl_int *err = NULL) const

Build info function that returns a vector of device/info pairs for the specified info type and for all devices in the program. On an error reading the info for any device, an empty vector of info will be returned.

inline cl_int createKernels(vector<Kernel> *kernels)

inline CL_API_PREFIX__VERSION_2_2_DEPRECATED cl_int setReleaseCallback (void(CL_CALLBACK *pfn_notify)(cl_program program, void *user_data), void *user_data=NULL) CL_API_SUFFIX__VERSION_2_2_DEPRECATED

Registers a callback function to be called when destructors for program scope global variables are complete and before the program is released.

Wraps clSetProgramReleaseCallback().

Each call to this function registers the specified user callback function on a callback stack associated with program. The registered user callback functions are called in the reverse order in which they were registered.

template<typename T>
inline std::enable_if<!std::is_pointer<T>::value, cl_int>::type setSpecializationConstant(cl_uint index, const T &value)

Sets a SPIR-V specialization constant.

Wraps clSetProgramSpecializationConstant().

inline cl_int setSpecializationConstant(cl_uint index, size_type size, const void *value)

Sets a SPIR-V specialization constant.

Wraps clSetProgramSpecializationConstant().

template<>
inline cl_int getInfo(cl_program_info name, vector<vector<unsigned char>> *param) const

template<>
inline vector<vector<unsigned char>> getInfo(cl_int *err) const

template<>
inline cl_int setSpecializationConstant(cl_uint index, const bool &value)

template<typename U>
struct rebind

#include <opencl.hpp>

Public Types

typedef SVMAllocator<U, SVMTrait> other

template<typename U>
struct rebind

#include <opencl.hpp>

Public Types

typedef SVMAllocator<U, SVMTrait> other

template<typename T>
struct ReferenceHandler

#include <opencl.hpp>

template<>
struct ReferenceHandler<cl_command_queue>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_command_queue queue)

static inline cl_int release(cl_command_queue queue)

template<>
struct ReferenceHandler<cl_context>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_context context)

static inline cl_int release(cl_context context)

template<>
struct ReferenceHandler<cl_device_id>

#include <opencl.hpp>

OpenCL 1.2 devices do have retain/release.

Public Static Functions

static inline cl_int retain(cl_device_id device)

Retain the device.

Parameters:

device – A valid device created using createSubDevices

Returns:

CL_SUCCESS if the function executed successfully. CL_INVALID_DEVICE if device was not a valid subdevice CL_OUT_OF_RESOURCES CL_OUT_OF_HOST_MEMORY

static inline cl_int release(cl_device_id device)

Retain the device.

Parameters:

device – A valid device created using createSubDevices

Returns:

CL_SUCCESS if the function executed successfully. CL_INVALID_DEVICE if device was not a valid subdevice CL_OUT_OF_RESOURCES CL_OUT_OF_HOST_MEMORY

template<>
struct ReferenceHandler<cl_event>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_event event)

static inline cl_int release(cl_event event)

template<>
struct ReferenceHandler<cl_kernel>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_kernel kernel)

static inline cl_int release(cl_kernel kernel)

template<>
struct ReferenceHandler<cl_mem>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_mem memory)

static inline cl_int release(cl_mem memory)

template<>
struct ReferenceHandler<cl_platform_id>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_platform_id)

static inline cl_int release(cl_platform_id)

template<>
struct ReferenceHandler<cl_program>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_program program)

static inline cl_int release(cl_program program)

template<>
struct ReferenceHandler<cl_sampler>

#include <opencl.hpp>

Public Static Functions

static inline cl_int retain(cl_sampler sampler)

static inline cl_int release(cl_sampler sampler)

template<typename ToType, typename CBRNG>
struct ReinterpretCtr

#include <ReinterpretCtr.hpp>

ReinterpretCtr uses memcpy to map back and forth between a CBRNG’s ctr_type and the specified ToType. For example, after:

typedef ReinterpretCtr<r123array4x32, Philox2x64> G;

G is a bona fide CBRNG with ctr_type r123array4x32.

WARNING: ReinterpretCtr is endian dependent. The values returned by G, declared as above, will depend on the endianness of the machine on which it runs.

Public Types

typedef ToType ctr_type

typedef CBRNG::key_type key_type

typedef CBRNG::ctr_type bctype

typedef CBRNG::ukey_type ukey_type

Public Functions

R123_STATIC_ASSERT (sizeof(ToType)==sizeof(bctype)&&sizeof(typename bctype::value_type)!=16,"ReinterpretCtr:  sizeof(ToType) is not the same as sizeof(CBRNG::ctr_type) or CBRNG::ctr_type::value_type looks like it might be __m128i")

inline ctr_type operator()(ctr_type c, key_type k)

class Sampler : public cl::detail::Wrapper<cl_sampler>

#include <opencl.hpp>

Class interface for cl_sampler.

See also

cl_sampler

Note

Copies of these objects are shallow, meaning that the copy will refer to the same underlying cl_sampler as the original. For details, see clRetainSampler() and clReleaseSampler().

Public Functions

inline Sampler()

Default constructor - initializes to NULL.

inline Sampler(const Context &context, cl_bool normalized_coords, cl_addressing_mode addressing_mode, cl_filter_mode filter_mode, cl_int *err = NULL)

Constructs a Sampler in a specified context.

Wraps clCreateSampler().

inline explicit Sampler(const cl_sampler &sampler, bool retainObject = false)

Constructor from cl_sampler - takes ownership.

Parameters:

retainObject – will cause the constructor to retain its cl object. Defaults to false to maintain compatibility with earlier versions. This effectively transfers ownership of a refcount on the cl_sampler into the new Sampler object.

inline Sampler &operator=(const cl_sampler &rhs)

Assignment operator from cl_sampler - takes ownership.

This effectively transfers ownership of a refcount on the rhs and calls clReleaseSampler() on the value previously held by this instance.

inline Sampler(const Sampler &sam)

Copy constructor to forward copy to the superclass correctly. Required for MSVC.

inline Sampler &operator=(const Sampler &sam)

Copy assignment to forward copy to the superclass correctly. Required for MSVC.

inline  Sampler (Sampler &&sam) CL_HPP_NOEXCEPT_

Move constructor to forward move to the superclass correctly. Required for MSVC.

inline Sampler &operator=(Sampler &&sam)

Move assignment to forward move to the superclass correctly. Required for MSVC.

template<typename T>
inline cl_int getInfo(cl_sampler_info name, T *param) const

Wrapper for clGetSamplerInfo().

template<cl_sampler_info name>
inline detail::param_traits<detail::cl_sampler_info, name>::param_type getInfo(cl_int *err = NULL) const

Wrapper for clGetSamplerInfo() that returns by value.

class SphericalAxes : public clsim::tabulator::Axes

#include <Axes.h>

[Half]-Spherical coordinate system centered on the source position, appropriate for approximately point-like sources

Public Functions

inline SphericalAxes(const std::vector<value_type> &axes)

virtual double GetBinVolume(const std::vector<size_t> &multiIndex) const

Calculate the volume (in m^3) of the cell at index multiIndex in the bin-content array.

Protected Functions

virtual std::string GetCoordinateFunction() const

Generate an OpenCL function that calculates the source-relative coordinates from the photon position and time.

virtual std::string GetBoundsCheckFunction() const

Generate an OpenCL function that returns true if the photon has exited the recording volume and should be stopped.

struct statistics_bundle

Public Functions

inline statistics_bundle()

Public Members

metric input_wait

metric output_wait

metric host_duration

metric device_duration

uint64_t total_host_duration

uint64_t total_device_duration

uint64_t total_queue_duration

uint64_t total_kernel_calls

uint64_t total_num_photons_generated

uint64_t total_num_photons_atDOMs

mutable boost::mutex mutex

struct statistics_bundle

Public Functions

inline statistics_bundle()

Public Members

metric input_wait

metric output_wait

metric host_duration

metric device_duration

uint64_t total_host_duration

uint64_t total_device_duration

uint64_t total_queue_duration

uint64_t total_kernel_calls

uint64_t total_num_photons_generated

uint64_t total_num_photons_atDOMs

mutable boost::mutex mutex

struct stringStruct

Public Members

int stringID

double meanX

double meanY

double maxZ

double minZ

double meandZ

double maxR

std::vector<domStruct> doms

unsigned short subdetectorNum

template<typename T, class SVMTrait>
class SVMAllocator

#include <opencl.hpp>

STL-like allocator class for managing SVM objects provided for convenience.

Note that while this behaves like an allocator for the purposes of constructing vectors and similar objects, care must be taken when using with smart pointers. The allocator should not be used to construct a unique_ptr if we are using coarse-grained SVM mode because the coarse-grained management behaviour would behave incorrectly with respect to reference counting.

Instead the allocator embeds a Deleter which may be used with unique_ptr and is used with the allocate_shared and allocate_ptr supplied operations.

Public Types

typedef T value_type

typedef value_type *pointer

typedef const value_type *const_pointer

typedef value_type &reference

typedef const value_type &const_reference

typedef std::size_t size_type

typedef std::ptrdiff_t difference_type

Public Functions

inline SVMAllocator()

inline explicit SVMAllocator(cl::Context context)

inline SVMAllocator(const SVMAllocator &other)

template<typename U>
inline SVMAllocator(const SVMAllocator<U, SVMTrait> &other)

inline ~SVMAllocator()

inline pointer address (reference r) CL_HPP_NOEXCEPT_

inline const_pointer address (const_reference r) CL_HPP_NOEXCEPT_

inline pointer allocate(size_type size, typename cl::SVMAllocator<void, SVMTrait>::const_pointer = 0)

Allocate an SVM pointer.

If the allocator is coarse-grained, this will take ownership to allow containers to correctly construct data in place.

inline void deallocate(pointer p, size_type)

inline size_type max_size () const CL_HPP_NOEXCEPT_

Return the maximum possible allocation size. This is the minimum of the maximum sizes of all devices in the context.

template<class U, class ...Args>
inline void construct(U *p, Args&&... args)

template<class U>
inline void destroy(U *p)

inline bool operator==(SVMAllocator const &rhs)

Returns true if the contexts match.

inline bool operator!=(SVMAllocator const &a)

Private Members

Context context_

Friends

friend class SVMAllocator

template<class SVMTrait>
class SVMAllocator<void, SVMTrait>

#include <opencl.hpp>

Public Types

typedef void value_type

typedef value_type *pointer

typedef const value_type *const_pointer

Friends

friend class SVMAllocator

template<class Trait = SVMTraitReadWrite<>>
class SVMTraitAtomic

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

template<class Trait = SVMTraitReadWrite<>>
class SVMTraitCoarse

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

template<class Trait = SVMTraitReadWrite<>>
class SVMTraitFine

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

class SVMTraitNull

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

template<class Trait = detail::SVMTraitNull>
class SVMTraitReadOnly

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

template<class Trait = detail::SVMTraitNull>
class SVMTraitReadWrite

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

template<class Trait = detail::SVMTraitNull>
class SVMTraitWriteOnly

#include <opencl.hpp>

Public Static Functions

static inline cl_svm_mem_flags getSVMMemFlags()

class TrkCerenkov : public G4VProcess

#include <TrkCerenkov.hh>

Public Functions

TrkCerenkov(const G4String &processName = "TrkCerenkov", G4ProcessType type = fElectromagnetic)

virtual ~TrkCerenkov()

inline G4bool IsApplicable(const G4ParticleDefinition &aParticleType)

G4double GetMeanFreePath(const G4Track &aTrack, G4double, G4ForceCondition*)

G4double PostStepGetPhysicalInteractionLength(const G4Track &aTrack, G4double, G4ForceCondition*)

G4VParticleChange *PostStepDoIt(const G4Track &aTrack, const G4Step &aStep)

inline virtual G4double AlongStepGetPhysicalInteractionLength(const G4Track&, G4double, G4double, G4double&, G4GPILSelection*)

inline virtual G4double AtRestGetPhysicalInteractionLength(const G4Track&, G4ForceCondition*)

inline virtual G4VParticleChange *AtRestDoIt(const G4Track&, const G4Step&)

inline virtual G4VParticleChange *AlongStepDoIt(const G4Track&, const G4Step&)

inline void SetMaxBetaChangePerStep(const G4double d)

inline void SetMaxNumPhotonsPerStep(const G4int NumPhotons)

inline void SetWlenBiasFunction(I3CLSimFunctionConstPtr wlenBias)

inline G4PhysicsTable *GetPhysicsTable() const

inline void DumpPhysicsTable() const

Protected Attributes

G4PhysicsTable *thePhysicsTable1

G4PhysicsTable *thePhysicsTable2

Private Functions

void BuildThePhysicsTable()

G4double GetAverageNumberOfPhotons(const G4double charge, const G4double beta, const G4Material *aMaterial, G4MaterialPropertyVector *Rindex) const

Private Members

G4double fMaxBetaChange

G4int fMaxPhotons

I3CLSimFunctionConstPtr fWlenBias

class TrkDetectorConstruction : public G4VUserDetectorConstruction

#include <TrkDetectorConstruction.hh>

Public Functions

TrkDetectorConstruction()

virtual ~TrkDetectorConstruction()

G4VPhysicalVolume *Construct()

G4Material *MaterialWithSingleIsotope(G4String, G4String, G4double, G4int, G4int)

void UpdateGeometry()

inline G4bool GetUpdated()

void SetMediumProperties(I3CLSimMediumPropertiesConstPtr)

Private Functions

void DefineMaterials()

G4VPhysicalVolume *ConstructDetector()

Private Members

G4bool updated

G4Box *world_sol

G4LogicalVolume *world_log

G4VPhysicalVolume *world_phys

G4Box *rock_sol

G4LogicalVolume *rock_log

G4VPhysicalVolume *rock_phys

G4Box *air_sol

G4LogicalVolume *air_log

G4VPhysicalVolume *air_phys

G4Element *O

G4Element *N

G4Element *C

G4Material *Air

G4Material *Vacuum

G4Material *H2O

G4Material *StdRock

G4Element *H

G4Element *Na

G4Element *Mg

G4Element *Cl

G4Element *Ca

G4Element *D

G4Element *O18

G4Material *normal_ice

G4Material *semiheavy_ice

G4Material *H2O18

G4Material *iso_ice

G4Material *Ice

G4MaterialPropertiesTable *MPT

I3CLSimMediumPropertiesConstPtr mediumProperties_

class TrkEventAction : public G4UserEventAction

#include <TrkEventAction.hh>

Public Functions

TrkEventAction()

virtual ~TrkEventAction()

void BeginOfEventAction(const G4Event*)

void EndOfEventAction(const G4Event*)

inline bool AbortWasRequested()

inline void SetExternalParticleID(uint32_t val)

inline void SetSecondaryCallback(const I3LightSourcePropagator::secondary_callback &callback)

inline void SetStepCallback(const I3LightSourcePropagator::step_callback &callback)

inline void SetMaxRefractiveIndex(double maxRefractiveIndex)

Private Members

bool abortRequested_

I3LightSourcePropagator::secondary_callback emitSecondary_

I3LightSourcePropagator::step_callback emitStep_

uint32_t currentExternalParticleID_

double maxRefractiveIndex_

class TrkOpticalPhysics : public G4VPhysicsConstructor

#include <TrkOpticalPhysics.hh>

Public Functions

TrkOpticalPhysics(const G4String &name, double maxBetaChangePerStep, uint32_t maxNumPhotonsPerStep)

virtual ~TrkOpticalPhysics()

virtual void ConstructParticle()

virtual void ConstructProcess()

void SetWlenBiasFunction(I3CLSimFunctionConstPtr wlenBias)

Protected Attributes

TrkCerenkov *theCerenkovProcess

Private Members

double maxBetaChangePerStep_

uint32_t maxNumPhotonsPerStep_

I3CLSimFunctionConstPtr wlenBias_

class TrkPrimaryGeneratorAction : public G4VUserPrimaryGeneratorAction

#include <TrkPrimaryGeneratorAction.hh>

Public Functions

TrkPrimaryGeneratorAction()

virtual ~TrkPrimaryGeneratorAction()

void GeneratePrimaries(G4Event *anEvent)

inline G4ParticleGun *GetParticleGun()

Private Members

G4ParticleGun *particleGun

class TrkStackingAction : public G4UserStackingAction

#include <TrkStackingAction.hh>

Public Functions

TrkStackingAction()

virtual ~TrkStackingAction()

virtual G4ClassificationOfNewTrack ClassifyNewTrack(const G4Track *aTrack)

virtual void NewStage()

virtual void PrepareNewEvent()

class TrkUISessionToQueue : public G4UIsession

#include <TrkUISessionToQueue.hh>

Public Functions

TrkUISessionToQueue(boost::shared_ptr<I3CLSimQueue<boost::shared_ptr<std::pair<const std::string, bool>>>> queueFromGeant4Messages)

virtual ~TrkUISessionToQueue()

virtual G4int ReceiveG4cout(G4String coutString)

virtual G4int ReceiveG4cerr(G4String cerrString)

Private Members

boost::shared_ptr<I3CLSimQueue<boost::shared_ptr<std::pair<const std::string, bool>>>> queueFromGeant4Messages_

class TrkUserEventInformation : public G4VUserEventInformation

#include <TrkUserEventInformation.hh>

Public Functions

TrkUserEventInformation(const I3LightSourcePropagator::secondary_callback &emitSecondary_, const I3LightSourcePropagator::step_callback &emitStep_, uint32_t currentExternalParticleID_, double maxRefractiveIndex_)

virtual ~TrkUserEventInformation()

inline void StartClock()

inline void StopClock()

inline double GetElapsedUserTime()

inline double GetElapsedSysTime()

inline double GetElapsedWallTime()

inline void Print() const

Public Members

bool abortRequested

const I3LightSourcePropagator::secondary_callback &emitSecondary

const I3LightSourcePropagator::step_callback &emitStep

uint32_t currentExternalParticleID

double maxRefractiveIndex

struct timeval start_wallclock_ end_wallclock_

struct rusage start_rusage_ end_rusage_

Private Functions

inline double delta(const struct timeval &starttime, const struct timeval &stoptime)

class UserEvent : public cl::Event

#include <opencl.hpp>

Class interface for user events (a subset of cl_event’s).

See Event for details about copy semantics, etc.

Public Functions

inline UserEvent(const Context &context, cl_int *err = NULL)

Constructs a user event on a given context.

Wraps clCreateUserEvent().

inline UserEvent()

Default constructor - initializes to NULL.

inline cl_int setStatus(cl_int status)

Sets the execution status of a user event object.

Wraps clSetUserEventStatus().

class UserHookForAbortState : public G4VStateDependent

Public Functions

inline UserHookForAbortState()

inline virtual ~UserHookForAbortState()

inline virtual G4bool Notify(G4ApplicationState requiredState)

template<typename T>
class Wrapper

#include <opencl.hpp>

Public Types

typedef T cl_type

Public Functions

inline Wrapper()

inline Wrapper(const cl_type &obj, bool retainObject)

inline ~Wrapper()

inline Wrapper(const Wrapper<cl_type> &rhs)

inline  Wrapper (Wrapper< cl_type > &&rhs) CL_HPP_NOEXCEPT_

inline Wrapper<cl_type> &operator=(const Wrapper<cl_type> &rhs)

inline Wrapper<cl_type> &operator=(Wrapper<cl_type> &&rhs)

inline Wrapper<cl_type> &operator=(const cl_type &rhs)

inline const cl_type &operator()() const

inline cl_type &operator()()

inline cl_type get() const

Protected Functions

inline cl_int retain() const

inline cl_int release() const

Protected Attributes

cl_type object_

Friends

template<typename Func, typename U>
inline friend cl_int getInfoHelper(Func, cl_uint, U*, int, typename U::cl_type)

template<>
class Wrapper<cl_device_id>

#include <opencl.hpp>

Subclassed by cl::Device

Public Types

typedef cl_device_id cl_type

Public Functions

inline Wrapper()

inline Wrapper(const cl_type &obj, bool retainObject)

inline ~Wrapper()

inline Wrapper(const Wrapper<cl_type> &rhs)

inline  Wrapper (Wrapper< cl_type > &&rhs) CL_HPP_NOEXCEPT_

inline Wrapper<cl_type> &operator=(const Wrapper<cl_type> &rhs)

inline Wrapper<cl_type> &operator=(Wrapper<cl_type> &&rhs)

inline Wrapper<cl_type> &operator=(const cl_type &rhs)

inline const cl_type &operator()() const

inline cl_type &operator()()

inline cl_type get() const

Protected Functions

inline cl_int retain() const

inline cl_int release() const

Protected Attributes

cl_type object_

bool referenceCountable_

Protected Static Functions

static inline bool isReferenceCountable(cl_device_id device)

Friends

template<typename Func, typename U>
inline friend cl_int getInfoHelper(Func, cl_uint, U*, int, typename U::cl_type)

template<typename Func, typename U>
inline friend cl_int getInfoHelper(Func, cl_uint, vector<U>*, int, typename U::cl_type)

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace cl

The OpenCL C++ bindings are defined within this namespace.

Typedefs

using size_type = size_t

using vector = std::vector<T, Alloc>

using string = std::string

using pointer = std::unique_ptr<T, D>

using array = std::array<T, N>

using BuildLogType = vector<std::pair<cl::Device, typename detail::param_traits<detail::cl_program_build_info, CL_PROGRAM_BUILD_LOG>::param_type>>

using coarse_svm_vector = vector<T, cl::SVMAllocator<int, cl::SVMTraitCoarse<>>>

Vector alias to simplify contruction of coarse-grained SVM containers.

using fine_svm_vector = vector<T, cl::SVMAllocator<int, cl::SVMTraitFine<>>>

Vector alias to simplify contruction of fine-grained SVM containers.

using atomic_svm_vector = vector<T, cl::SVMAllocator<int, cl::SVMTraitAtomic<>>>

Vector alias to simplify contruction of fine-grained SVM containers that support platform atomics.

Enums

enum QueueProperties

Values:

enumerator None

enumerator Profiling

enumerator OutOfOrder

enum DeviceQueueProperties

Values:

enumerator None

enumerator Profiling

Functions

static inline cl_int WaitForEvents(const vector<Event> &events)

Blocks the calling thread until every event specified is complete.

Wraps clWaitForEvents().

template<typename IteratorType>
inline cl_int copy(IteratorType startIterator, IteratorType endIterator, cl::Buffer &buffer)

Blocking copy operation between iterators and a buffer. Host to Device. Uses default command queue.

template<typename IteratorType>
inline cl_int copy(const cl::Buffer &buffer, IteratorType startIterator, IteratorType endIterator)

Blocking copy operation between iterators and a buffer. Device to Host. Uses default command queue.

template<typename IteratorType>
inline cl_int copy(const CommandQueue &queue, IteratorType startIterator, IteratorType endIterator, cl::Buffer &buffer)

Blocking copy operation between iterators and a buffer. Host to Device. Uses specified queue.

template<typename IteratorType>
inline cl_int copy(const CommandQueue &queue, const cl::Buffer &buffer, IteratorType startIterator, IteratorType endIterator)

Blocking copy operation between iterators and a buffer. Device to Host. Uses specified queue.

template<typename T>
inline cl_int enqueueMapSVM(T *ptr, cl_bool blocking, cl_map_flags flags, size_type size, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a raw SVM pointer.

template<class T, class Alloc, class ...Args>
cl::pointer<T, detail::Deleter<Alloc>> allocate_pointer(const Alloc &alloc_, Args&&... args)

Allocation operation compatible with std::allocate_ptr. Creates a unique_ptr<T> by default. This requirement is to ensure that the control block is not allocated in memory inaccessible to the host.

template<class T, class SVMTrait, class ...Args>
cl::pointer<T, detail::Deleter<SVMAllocator<T, SVMTrait>>> allocate_svm(Args... args)

template<class T, class SVMTrait, class ...Args>
cl::pointer<T, detail::Deleter<SVMAllocator<T, SVMTrait>>> allocate_svm(const cl::Context &c, Args... args)

inline LocalSpaceArg Local(size_type size)

Helper function for generating LocalSpaceArg objects.

Local

inline Program linkProgram (Program input1, Program input2, const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL, cl_int *err=NULL)

inline Program linkProgram (vector< Program > inputPrograms, const char *options=NULL, void(CL_CALLBACK *notifyFptr)(cl_program, void *)=NULL, void *data=NULL, cl_int *err=NULL)

inline QueueProperties operator|(QueueProperties lhs, QueueProperties rhs)

inline DeviceQueueProperties operator|(DeviceQueueProperties lhs, DeviceQueueProperties rhs)

inline cl_int enqueueReadBuffer(const Buffer &buffer, cl_bool blocking, size_type offset, size_type size, void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueWriteBuffer(const Buffer &buffer, cl_bool blocking, size_type offset, size_type size, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline void *enqueueMapBuffer(const Buffer &buffer, cl_bool blocking, cl_map_flags flags, size_type offset, size_type size, const vector<Event> *events = NULL, Event *event = NULL, cl_int *err = NULL)

template<typename T, class D>
inline cl_int enqueueMapSVM(cl::pointer<T, D> ptr, cl_bool blocking, cl_map_flags flags, size_type size, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a cl::pointer instance.

template<typename T, class Alloc>
inline cl_int enqueueMapSVM(cl::vector<T, Alloc> container, cl_bool blocking, cl_map_flags flags, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will allow the host to update a region of a coarse-grained SVM buffer. This variant takes a cl::vector instance.

inline cl_int enqueueUnmapMemObject(const Memory &memory, void *mapped_ptr, const vector<Event> *events = NULL, Event *event = NULL)

template<typename T>
inline cl_int enqueueUnmapSVM(T *ptr, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a raw SVM pointer.

template<typename T, class D>
inline cl_int enqueueUnmapSVM(cl::pointer<T, D> &ptr, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a cl::pointer instance.

template<typename T, class Alloc>
inline cl_int enqueueUnmapSVM(cl::vector<T, Alloc> &container, const vector<Event> *events = NULL, Event *event = NULL)

Enqueues to the default queue a command that will release a coarse-grained SVM buffer back to the OpenCL runtime. This variant takes a cl::vector instance.

inline cl_int enqueueCopyBuffer(const Buffer &src, const Buffer &dst, size_type src_offset, size_type dst_offset, size_type size, const vector<Event> *events = NULL, Event *event = NULL)

template<typename T, class Alloc>
inline cl_int mapSVM(cl::vector<T, Alloc> &container)

Blocking SVM map operation - performs a blocking map underneath.

template<typename T, class Alloc>
inline cl_int unmapSVM(cl::vector<T, Alloc> &container)

Blocking SVM map operation - performs a blocking map underneath.

inline cl_int enqueueReadBufferRect(const Buffer &buffer, cl_bool blocking, const array<size_type, 3> &buffer_offset, const array<size_type, 3> &host_offset, const array<size_type, 3> &region, size_type buffer_row_pitch, size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueWriteBufferRect(const Buffer &buffer, cl_bool blocking, const array<size_type, 3> &buffer_offset, const array<size_type, 3> &host_offset, const array<size_type, 3> &region, size_type buffer_row_pitch, size_type buffer_slice_pitch, size_type host_row_pitch, size_type host_slice_pitch, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueCopyBufferRect(const Buffer &src, const Buffer &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, size_type src_row_pitch, size_type src_slice_pitch, size_type dst_row_pitch, size_type dst_slice_pitch, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueReadImage(const Image &image, cl_bool blocking, const array<size_type, 3> &origin, const array<size_type, 3> &region, size_type row_pitch, size_type slice_pitch, void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueWriteImage(const Image &image, cl_bool blocking, const array<size_type, 3> &origin, const array<size_type, 3> &region, size_type row_pitch, size_type slice_pitch, const void *ptr, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueCopyImage(const Image &src, const Image &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueCopyImageToBuffer(const Image &src, const Buffer &dst, const array<size_type, 3> &src_origin, const array<size_type, 3> &region, size_type dst_offset, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int enqueueCopyBufferToImage(const Buffer &src, const Image &dst, size_type src_offset, const array<size_type, 3> &dst_origin, const array<size_type, 3> &region, const vector<Event> *events = NULL, Event *event = NULL)

inline cl_int flush(void)

inline cl_int finish(void)

Variables

static const NDRange NullRange

A zero-dimensional range.

namespace compatibility

namespace detail

Typedefs

using size_t_array = array<size_type, 3>

Functions

static inline cl_int errHandler(cl_int err, const char *errStr = NULL)

template<typename Functor, typename T>
inline cl_int getInfoHelper(Functor f, cl_uint name, T *param, long)

template<typename Func>
inline cl_int getInfoHelper(Func f, cl_uint name, vector<vector<unsigned char>> *param, int)

template<typename Func, typename T>
inline cl_int getInfoHelper(Func f, cl_uint name, vector<T> *param, long)

template<typename Func, typename T>
inline cl_int getInfoHelper(Func f, cl_uint name, vector<T> *param, int, typename T::cl_type = 0)

template<typename Func>
inline cl_int getInfoHelper(Func f, cl_uint name, string *param, long)

template<typename Func, size_type N>
inline cl_int getInfoHelper(Func f, cl_uint name, array<size_type, N> *param, long)

template<typename Func, typename T>
inline cl_int getInfoHelper(Func f, cl_uint name, T *param, int, typename T::cl_type = 0)

template<typename Func, typename T>
inline cl_int getInfo(Func f, cl_uint name, T *param)

template<typename Func, typename Arg0, typename T>
inline cl_int getInfo(Func f, const Arg0 &arg0, cl_uint name, T *param)

template<typename Func, typename Arg0, typename Arg1, typename T>
inline cl_int getInfo(Func f, const Arg0 &arg0, const Arg1 &arg1, cl_uint name, T *param)

template<typename T>
inline bool operator==(const Wrapper<T> &lhs, const Wrapper<T> &rhs)

template<typename T>
inline bool operator!=(const Wrapper<T> &lhs, const Wrapper<T> &rhs)

static inline cl_int buildErrHandler(cl_int err, const char *errStr, const BuildLogType &buildLogs)

namespace clsim

namespace tabulator

Functions

I3_POINTER_TYPEDEFS(Axes)

namespace I3CLSimHelper

Functions

bool write_geometry_code_and_fill_buffer(std::string &code, const std::vector<int> &stringIDs, const std::vector<unsigned int> &domIDs, const std::vector<double> &posX, const std::vector<double> &posY, const std::vector<double> &posZ, const std::vector<double> &cableAngle, const std::vector<std::string> &subdetectors, const double omRadius, const double cableRadius, std::vector<cl_ushort> &geoLayerToOMNumIndexPerStringSetBuffer, std::vector<int> &stringIndexToStringIDBuffer, std::vector<std::vector<unsigned int>> &domIndexToDomIDBuffer_perStringIndex)

std::string GenerateGeometrySource(const I3SimpleGeometry &geometry, std::vector<unsigned short> &geoLayerToOMNumIndexPerStringSetBuffer, std::vector<int> &stringIndexToStringIDBuffer, std::vector<std::vector<unsigned int>> &domIndexToDomIDBuffer_perStringIndex)

generates the OpenCL source code for a given I3SimpleGeometry object.

bool divideIntoCells(const std::vector<stringStruct> &strings, int subdetectorNum, double &cellStartX, double &cellStartY, double &cellWidthX, double &cellWidthY, unsigned int cellNumX, unsigned int cellNumY, std::vector<unsigned short> &cellToStringIndex)

bool doesMatchLayering(const stringStruct &currentString, double layerStartZ, double layerHeight, unsigned int layerNum, const double domRadius, const std::vector<unsigned short> &layerToOMNumIndex)

void findOverallStringMinMaxZ(const std::vector<stringStruct> &strings, const double domRadius, double &minZ, double &maxZ, bool includeDomRadius)

bool divideIntoLayers(const stringStruct &currentString, double &layerStartZ, double &layerHeight, unsigned int layerNum, const double domRadius, double minZHint, double maxZHint, std::vector<unsigned short> &layerToOMNumIndex)

std::string generate_get_dom_position_code(const std::vector<stringStruct> &strings, bool includeCable)

template<class T>
bool is_in_vector(const T &compVal, const std::vector<T> &vect, std::size_t &pos)

template<class T>
void OptimizeLayeredValue(const std::vector<boost::shared_ptr<const T>> &inputValues, std::vector<boost::shared_ptr<const T>> &outputFunctions, std::vector<std::size_t> &outputFunctionForLayer)

std::string GenerateLayeredWlenDependentFunctions_WriteMainFuncCode(const std::vector<std::size_t> &outputFunctionForLayer, const std::vector<I3CLSimFunctionConstPtr> &functionsToGenerate, const std::string &fullName, const std::string &functionName)

std::string GenerateLayeredWlenDependentFunctions(const std::vector<I3CLSimFunctionConstPtr> &layeredFunction, const std::string &fullName, const std::string &functionName, std::string derivativeFunctionName = "")

std::string GenerateMediumPropertiesSource(const I3CLSimMediumProperties &mediumProperties)

generates the OpenCL source code for a given mediumProperties object.

std::string GenerateWavelengthGeneratorSource(const std::vector<I3CLSimRandomValueConstPtr> &wlenGenerators)

template<class T, class U>
std::vector<boost::shared_ptr<const T>> dynamic_vector_contents_cast(const std::vector<boost::shared_ptr<const U>> &inVec, bool &worked)

template<typename T, class V>
bool is_constant(const std::vector<T> &vect, const V &visitor)

std::string GenerateOptimizedCodeFor_I3CLSimFunctionAbsLenIceCube(const std::vector<I3CLSimFunctionConstPtr> &layeredFunction, const std::string &fullName, const std::string &functionName, bool &worked)

std::string GenerateOptimizedCodeFor_I3CLSimFunctionScatLenIceCube(const std::vector<I3CLSimFunctionConstPtr> &layeredFunction, const std::string &fullName, const std::string &functionName, bool &worked)

inline std::string LoadProgramSource(const std::string &filename)

std::string GetMathPreamble(const I3CLSimOpenCLDevice &device, bool double_precision = false)

namespace [anonymous]

namespace [anonymous]

namespace [anonymous]

namespace I3CLSimI3ParticleGeantConverter

Helper function: Converts I3Particles from/to G4Particles.

It either takes an G4Track object and sets the properties of an I3Particle to correspond to it takes an I3Particle and sets the properties of an G4ParticleGun to shoot paticles of this type.

Functions

inline I3Particle::ParticleType ConvertPDGEncodingToI3ParticleType(int pdg_code)

inline int ConvertI3ParticleTypeToPDGEncoding(I3Particle::ParticleType type)

inline bool SetParticleGun(G4ParticleGun *particleGun, const I3Particle &particle)

Configures a G4ParticleGun to shoot particles as configured by an I3Particle. Returns “false” if the I3Particle does not contain enough information (i.e., some of the required fields are NaN).

namespace I3CLSimLightSourceToStepConverterUtils

Functions

static double f(double energy, void *params)

double NumberOfPhotonsPerMeter(const I3CLSimFunction &phaseRefIndex, const I3CLSimFunction &wavelengthGenerationBias, double fromWlen, double toWlen)

static double f2(double wlen, void *params)

double PhotonNumberCorrectionFactorAfterBias(const I3CLSimFunction &unbiasedSpectrum, const I3CLSimFunction &wavelengthGenerationBias, double fromWlen, double toWlen)

inline uint64_t mwcRngInitState(I3RandomServicePtr randomService, uint32_t a)

inline double mwcRngRandomNumber_co(uint64_t &state, uint32_t a)

inline double mwcRngRandomNumber_oc(uint64_t &state, uint32_t a)

inline double gammaDistributedNumber(double shape, uint64_t &rngState, uint32_t rngA)

inline double gammaDistributedNumber(double shape, I3RandomService &randomService)

inline double gammaDistributedNumber(double shape, I3RandomServicePtr randomService)

inline void scatterDirectionByAngle(double cosa, double sina, double &x, double &y, double &z, double randomValue)

namespace I3CLSimModuleHelper

Functions

I3CLSimRandomValueConstPtr makeWavelengthGenerator(I3CLSimFunctionConstPtr unbiasedSpectrum, I3CLSimFunctionConstPtr wavelengthGenerationBias, I3CLSimMediumPropertiesConstPtr mediumProperties)

I3CLSimRandomValueConstPtr makeCherenkovWavelengthGenerator(I3CLSimFunctionConstPtr wavelengthGenerationBias, bool generateCherenkovPhotonsWithoutDispersion, I3CLSimMediumPropertiesConstPtr mediumProperties)

namespace [anonymous]

namespace I3ShadowPhotonRemover

This module removes photons that have paths intersecting with any shadowing part of the detecor (such as cables). This code is NOT functional at the moment.

Functions

bool is_shadowed(const I3Position &photon_position, const I3Direction &photon_direction, const I3ExtraGeometryItemCylinder &cylinder, const double distance)

Given a cylinder and a photon determine whether the photon intersects the cylinder backpropagating a given distance.

namespace r123

Most of the Random123 C++ API is contained in the r123 namespace.

namespace util

namespace std

STL namespace.

file aes.h

#include “features/compilerfeatures.h”

#include “array.h”

file array.h

#include “features/compilerfeatures.h”

#include “features/sse.h”

Defines

CXXMETHODS(_N, W, T)

CXXOVERLOADS(_N, W, T)

_r123array_tpl(_N, W, T)

R123_W(a)

Functions

_r123array_tpl (1, 32, uint32_t) _r123array_tpl(2

uint32_t _r123array_tpl (4, 32, uint32_t) _r123array_tpl(8

uint32_t uint32_t _r123array_tpl (1, 64, uint64_t) _r123array_tpl(2

uint32_t uint32_t uint64_t _r123array_tpl (4, 64, uint64_t) _r123array_tpl(16

file ars.h

#include “features/compilerfeatures.h”

#include “array.h”

file Axes.cxx

#include “icetray/I3Units.h”

#include “clsim/tabulator/Axes.h”

#include “simclasses/util/ToFloatString.h”

#include “opencl/I3CLSimHelperLoadProgramSource.h”

#include <boost/foreach.hpp>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

file Axes.h

#include “icetray/I3PointerTypedefs.h”

#include “clsim/tabulator/Axis.h”

#include <vector>

#include <string>

#include <boost/noncopyable.hpp>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

file Axis.cxx

#include “clsim/tabulator/Axis.h”

#include “simclasses/util/ToFloatString.h”

#include <boost/foreach.hpp>

#include <cmath>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

file Axis.h

#include <vector>

#include <string>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

file cl.hpp

#include “clsim/opencl.hpp”

Defines

CL_HPP_ENABLE_EXCEPTIONS

CL_HPP_TARGET_OPENCL_VERSION

CL_HPP_MINIMUM_OPENCL_VERSION

file clangfeatures.h

#include “gccfeatures.h”

Defines

R123_USE_CXX11_UNRESTRICTED_UNIONS

R123_USE_CXX11_STATIC_ASSERT

R123_STATIC_ASSERT(expr, msg)

R123_USE_CXX11_CONSTEXPR

R123_USE_CXX11_EXPLICIT_CONVERSIONS

R123_USE_CXX11_RANDOM

R123_USE_CXX11_TYPE_TRAITS

file compilerfeatures.h

file convert.cxx

#include <icetray/open.h>

#include <icetray/I3Logging.h>

Functions

int main(int argc, char const *argv[])

file Engine.hpp

#include “../features/compilerfeatures.h”

#include “../array.h”

#include <limits>

#include <stdexcept>

#include <sstream>

#include <algorithm>

#include <vector>

file gccfeatures.h

#include <including_a_nonexistent_file_will_stop_some_compilers_from_continuing_with_a_hopeless_task>

#include <assert.h>

#include <stdint.h>

Defines

R123_GNUC_VERSION

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_CXX11_UNRESTRICTED_UNIONS

R123_USE_CXX11_STATIC_ASSERT

R123_USE_CXX11_CONSTEXPR

R123_USE_CXX11_EXPLICIT_CONVERSIONS

R123_USE_CXX11_RANDOM

R123_USE_CXX11_TYPE_TRAITS

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_CPUID_MSVC

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

R123_MULHILO64_MULHI_INTRIN

R123_USE_MULHILO32_MULHI_INTRIN

R123_MULHILO32_MULHI_INTRIN

__STDC_CONSTANT_MACROS

file gsl_cbrng.h

#include <gsl/gsl_rng.h>

#include <string.h>

Defines

GSL_CBRNG(NAME, CBRNGNAME)

The macro: GSL_CBRNG(NAME, CBRNGNAME) declares the necessary structs and constants that define a gsl_rng_NAME type based on the counter-based RNG CBRNGNAME. For example:

Usage:

#include <Random123/threefry.h>
#include <Random123/conventional/gsl_cbrng.h>  // this file
GSL_CBRNG(cbrng, threefry4x32); // creates gsl_rng_cbrng

int main(int argc, char **argv){
    gsl_rng *r = gsl_rng_alloc(gsl_rng_cbrng);
    ... use r as you would use any other gsl_rng ...   
 }

It requires that NAME be the name of a CBRNG that follows the naming and stylistic conventions of the Random123 library.

Note that wrapping a counter-based PRNG with a traditional API in this way obscures much of the power of the CBRNG API. Nevertheless, it may be of value to applications that are already coded to work with GSL random number generators, and that wish to use the RNGs in the Random123 library.

file gsl_microrng.h

#include <gsl/gsl_rng.h>

#include <string.h>

Defines

GSL_MICRORNG(NAME, CBRNGNAME)

The macro: GSL_MICRORNG(NAME, CBRNGNAME) is the GSL analog analog of the C++ r123::MicroURNG template. It declares a gsl_rng type named gsl_rng_NAME which uses the underlying CBRNGNAME and can be invoked a limited number of times between calls to NAME_reset.

When the underlying CBRNG’s ctr_t is an r123arrayNxW, and the gsl_rng_NAME may called up to N*2^32 times between calls to NAME_reset.

NAME_reset takes a gsl_rng_NAME type, a counter and a key as arguments. It restarts the micro-rng with a new base counter and key.

Note that you must call NAME_reset before the first use of a gsl_rng. NAME_reset is not called automatically by gsl_rng_alloc().

#include <Random123/threefry.h>
#include <Random123/gsl_microrng.h> // this file
GSL_MICRORNG(microcbrng, threefry4x64, 20)      // creates gsl_rng_microcbrng

int main(int argc, char** argv) {
     gsl_rng *r = gsl_rng_alloc(gsl_rng_microcbrng);
     threefry4x64_ctr_t c = {{}};
     threefry4x64_key_t k = {{}};

     for (...) {
         c.v[0] = ??; //  some application variable
         microcbrng_reset(r, c, k);
         for (...) {
        // gaussian calls r several times.  It is safe for
        // r to be used upto 2^20 times in this loop
        something[i] = gsl_ran_gaussian(r, 1.5);
         }
     }
}

file helper_math.h

#include “cuda_runtime.h”

#include <math.h>

Defines

EXIT_WAIVED

Typedefs

typedef unsigned int uint

typedef unsigned short ushort

Functions

inline float fminf(float a, float b)

inline float fmaxf(float a, float b)

inline int max(int a, int b)

inline int min(int a, int b)

inline float rsqrtf(float x)

inline __host__ __device__ float2 make_float2 (float s)

inline __host__ __device__ float2 make_float2 (float3 a)

inline __host__ __device__ float2 make_float2 (int2 a)

inline __host__ __device__ float2 make_float2 (uint2 a)

inline __host__ __device__ int2 make_int2 (int s)

inline __host__ __device__ int2 make_int2 (int3 a)

inline __host__ __device__ int2 make_int2 (uint2 a)

inline __host__ __device__ int2 make_int2 (float2 a)

inline __host__ __device__ uint2 make_uint2 (uint s)

inline __host__ __device__ uint2 make_uint2 (uint3 a)

inline __host__ __device__ uint2 make_uint2 (int2 a)

inline __host__ __device__ float3 make_float3 (float s)

inline __host__ __device__ float3 make_float3 (float2 a)

inline __host__ __device__ float3 make_float3 (float2 a, float s)

inline __host__ __device__ float3 make_float3 (float4 a)

inline __host__ __device__ float3 make_float3 (int3 a)

inline __host__ __device__ float3 make_float3 (uint3 a)

inline __host__ __device__ int3 make_int3 (int s)

inline __host__ __device__ int3 make_int3 (int2 a)

inline __host__ __device__ int3 make_int3 (int2 a, int s)

inline __host__ __device__ int3 make_int3 (uint3 a)

inline __host__ __device__ int3 make_int3 (float3 a)

inline __host__ __device__ uint3 make_uint3 (uint s)

inline __host__ __device__ uint3 make_uint3 (uint2 a)

inline __host__ __device__ uint3 make_uint3 (uint2 a, uint s)

inline __host__ __device__ uint3 make_uint3 (uint4 a)

inline __host__ __device__ uint3 make_uint3 (int3 a)

inline __host__ __device__ float4 make_float4 (float s)

inline __host__ __device__ float4 make_float4 (float3 a)

inline __host__ __device__ float4 make_float4 (float3 a, float w)

inline __host__ __device__ float4 make_float4 (int4 a)

inline __host__ __device__ float4 make_float4 (uint4 a)

inline __host__ __device__ int4 make_int4 (int s)

inline __host__ __device__ int4 make_int4 (int3 a)

inline __host__ __device__ int4 make_int4 (int3 a, int w)

inline __host__ __device__ int4 make_int4 (uint4 a)

inline __host__ __device__ int4 make_int4 (float4 a)

inline __host__ __device__ uint4 make_uint4 (uint s)

inline __host__ __device__ uint4 make_uint4 (uint3 a)

inline __host__ __device__ uint4 make_uint4 (uint3 a, uint w)

inline __host__ __device__ uint4 make_uint4 (int4 a)

inline __host__ __device__ float2 operator- (float2 &a)

inline __host__ __device__ int2 operator- (int2 &a)

inline __host__ __device__ float3 operator- (float3 &a)

inline __host__ __device__ int3 operator- (int3 &a)

inline __host__ __device__ float4 operator- (float4 &a)

inline __host__ __device__ int4 operator- (int4 &a)

inline __host__ __device__ float2 operator+ (float2 a, float2 b)

inline __host__ __device__ void operator+= (float2 &a, float2 b)

inline __host__ __device__ float2 operator+ (float2 a, float b)

inline __host__ __device__ float2 operator+ (float b, float2 a)

inline __host__ __device__ void operator+= (float2 &a, float b)

inline __host__ __device__ int2 operator+ (int2 a, int2 b)

inline __host__ __device__ void operator+= (int2 &a, int2 b)

inline __host__ __device__ int2 operator+ (int2 a, int b)

inline __host__ __device__ int2 operator+ (int b, int2 a)

inline __host__ __device__ void operator+= (int2 &a, int b)

inline __host__ __device__ uint2 operator+ (uint2 a, uint2 b)

inline __host__ __device__ void operator+= (uint2 &a, uint2 b)

inline __host__ __device__ uint2 operator+ (uint2 a, uint b)

inline __host__ __device__ uint2 operator+ (uint b, uint2 a)

inline __host__ __device__ void operator+= (uint2 &a, uint b)

inline __host__ __device__ float3 operator+ (float3 a, float3 b)

inline __host__ __device__ void operator+= (float3 &a, float3 b)

inline __host__ __device__ float3 operator+ (float3 a, float b)

inline __host__ __device__ void operator+= (float3 &a, float b)

inline __host__ __device__ int3 operator+ (int3 a, int3 b)

inline __host__ __device__ void operator+= (int3 &a, int3 b)

inline __host__ __device__ int3 operator+ (int3 a, int b)

inline __host__ __device__ void operator+= (int3 &a, int b)

inline __host__ __device__ uint3 operator+ (uint3 a, uint3 b)

inline __host__ __device__ void operator+= (uint3 &a, uint3 b)

inline __host__ __device__ uint3 operator+ (uint3 a, uint b)

inline __host__ __device__ void operator+= (uint3 &a, uint b)

inline __host__ __device__ int3 operator+ (int b, int3 a)

inline __host__ __device__ uint3 operator+ (uint b, uint3 a)

inline __host__ __device__ float3 operator+ (float b, float3 a)

inline __host__ __device__ float4 operator+ (float4 a, float4 b)

inline __host__ __device__ void operator+= (float4 &a, float4 b)

inline __host__ __device__ float4 operator+ (float4 a, float b)

inline __host__ __device__ float4 operator+ (float b, float4 a)

inline __host__ __device__ void operator+= (float4 &a, float b)

inline __host__ __device__ int4 operator+ (int4 a, int4 b)

inline __host__ __device__ void operator+= (int4 &a, int4 b)

inline __host__ __device__ int4 operator+ (int4 a, int b)

inline __host__ __device__ int4 operator+ (int b, int4 a)

inline __host__ __device__ void operator+= (int4 &a, int b)

inline __host__ __device__ uint4 operator+ (uint4 a, uint4 b)

inline __host__ __device__ void operator+= (uint4 &a, uint4 b)

inline __host__ __device__ uint4 operator+ (uint4 a, uint b)

inline __host__ __device__ uint4 operator+ (uint b, uint4 a)

inline __host__ __device__ void operator+= (uint4 &a, uint b)

inline __host__ __device__ float2 operator- (float2 a, float2 b)

inline __host__ __device__ void operator-= (float2 &a, float2 b)

inline __host__ __device__ float2 operator- (float2 a, float b)

inline __host__ __device__ float2 operator- (float b, float2 a)

inline __host__ __device__ void operator-= (float2 &a, float b)

inline __host__ __device__ int2 operator- (int2 a, int2 b)

inline __host__ __device__ void operator-= (int2 &a, int2 b)

inline __host__ __device__ int2 operator- (int2 a, int b)

inline __host__ __device__ int2 operator- (int b, int2 a)

inline __host__ __device__ void operator-= (int2 &a, int b)

inline __host__ __device__ uint2 operator- (uint2 a, uint2 b)

inline __host__ __device__ void operator-= (uint2 &a, uint2 b)

inline __host__ __device__ uint2 operator- (uint2 a, uint b)

inline __host__ __device__ uint2 operator- (uint b, uint2 a)

inline __host__ __device__ void operator-= (uint2 &a, uint b)

inline __host__ __device__ float3 operator- (float3 a, float3 b)

inline __host__ __device__ void operator-= (float3 &a, float3 b)

inline __host__ __device__ float3 operator- (float3 a, float b)

inline __host__ __device__ float3 operator- (float b, float3 a)

inline __host__ __device__ void operator-= (float3 &a, float b)

inline __host__ __device__ int3 operator- (int3 a, int3 b)

inline __host__ __device__ void operator-= (int3 &a, int3 b)

inline __host__ __device__ int3 operator- (int3 a, int b)

inline __host__ __device__ int3 operator- (int b, int3 a)

inline __host__ __device__ void operator-= (int3 &a, int b)

inline __host__ __device__ uint3 operator- (uint3 a, uint3 b)

inline __host__ __device__ void operator-= (uint3 &a, uint3 b)

inline __host__ __device__ uint3 operator- (uint3 a, uint b)

inline __host__ __device__ uint3 operator- (uint b, uint3 a)

inline __host__ __device__ void operator-= (uint3 &a, uint b)

inline __host__ __device__ float4 operator- (float4 a, float4 b)

inline __host__ __device__ void operator-= (float4 &a, float4 b)

inline __host__ __device__ float4 operator- (float4 a, float b)

inline __host__ __device__ void operator-= (float4 &a, float b)

inline __host__ __device__ int4 operator- (int4 a, int4 b)

inline __host__ __device__ void operator-= (int4 &a, int4 b)

inline __host__ __device__ int4 operator- (int4 a, int b)

inline __host__ __device__ int4 operator- (int b, int4 a)

inline __host__ __device__ void operator-= (int4 &a, int b)

inline __host__ __device__ uint4 operator- (uint4 a, uint4 b)

inline __host__ __device__ void operator-= (uint4 &a, uint4 b)

inline __host__ __device__ uint4 operator- (uint4 a, uint b)

inline __host__ __device__ uint4 operator- (uint b, uint4 a)

inline __host__ __device__ void operator-= (uint4 &a, uint b)

inline __host__ __device__ float2 operator* (float2 a, float2 b)

inline __host__ __device__ void operator*= (float2 &a, float2 b)

inline __host__ __device__ float2 operator* (float2 a, float b)

inline __host__ __device__ float2 operator* (float b, float2 a)

inline __host__ __device__ void operator*= (float2 &a, float b)

inline __host__ __device__ int2 operator* (int2 a, int2 b)

inline __host__ __device__ void operator*= (int2 &a, int2 b)

inline __host__ __device__ int2 operator* (int2 a, int b)

inline __host__ __device__ int2 operator* (int b, int2 a)

inline __host__ __device__ void operator*= (int2 &a, int b)

inline __host__ __device__ uint2 operator* (uint2 a, uint2 b)

inline __host__ __device__ void operator*= (uint2 &a, uint2 b)

inline __host__ __device__ uint2 operator* (uint2 a, uint b)

inline __host__ __device__ uint2 operator* (uint b, uint2 a)

inline __host__ __device__ void operator*= (uint2 &a, uint b)

inline __host__ __device__ float3 operator* (float3 a, float3 b)

inline __host__ __device__ void operator*= (float3 &a, float3 b)

inline __host__ __device__ float3 operator* (float3 a, float b)

inline __host__ __device__ float3 operator* (float b, float3 a)

inline __host__ __device__ void operator*= (float3 &a, float b)

inline __host__ __device__ int3 operator* (int3 a, int3 b)

inline __host__ __device__ void operator*= (int3 &a, int3 b)

inline __host__ __device__ int3 operator* (int3 a, int b)

inline __host__ __device__ int3 operator* (int b, int3 a)

inline __host__ __device__ void operator*= (int3 &a, int b)

inline __host__ __device__ uint3 operator* (uint3 a, uint3 b)

inline __host__ __device__ void operator*= (uint3 &a, uint3 b)

inline __host__ __device__ uint3 operator* (uint3 a, uint b)

inline __host__ __device__ uint3 operator* (uint b, uint3 a)

inline __host__ __device__ void operator*= (uint3 &a, uint b)

inline __host__ __device__ float4 operator* (float4 a, float4 b)

inline __host__ __device__ void operator*= (float4 &a, float4 b)

inline __host__ __device__ float4 operator* (float4 a, float b)

inline __host__ __device__ float4 operator* (float b, float4 a)

inline __host__ __device__ void operator*= (float4 &a, float b)

inline __host__ __device__ int4 operator* (int4 a, int4 b)

inline __host__ __device__ void operator*= (int4 &a, int4 b)

inline __host__ __device__ int4 operator* (int4 a, int b)

inline __host__ __device__ int4 operator* (int b, int4 a)

inline __host__ __device__ void operator*= (int4 &a, int b)

inline __host__ __device__ uint4 operator* (uint4 a, uint4 b)

inline __host__ __device__ void operator*= (uint4 &a, uint4 b)

inline __host__ __device__ uint4 operator* (uint4 a, uint b)

inline __host__ __device__ uint4 operator* (uint b, uint4 a)

inline __host__ __device__ void operator*= (uint4 &a, uint b)

inline __host__ __device__ float2 operator/ (float2 a, float2 b)

inline __host__ __device__ void operator/= (float2 &a, float2 b)

inline __host__ __device__ float2 operator/ (float2 a, float b)

inline __host__ __device__ void operator/= (float2 &a, float b)

inline __host__ __device__ float2 operator/ (float b, float2 a)

inline __host__ __device__ float3 operator/ (float3 a, float3 b)

inline __host__ __device__ void operator/= (float3 &a, float3 b)

inline __host__ __device__ float3 operator/ (float3 a, float b)

inline __host__ __device__ void operator/= (float3 &a, float b)

inline __host__ __device__ float3 operator/ (float b, float3 a)

inline __host__ __device__ float4 operator/ (float4 a, float4 b)

inline __host__ __device__ void operator/= (float4 &a, float4 b)

inline __host__ __device__ float4 operator/ (float4 a, float b)

inline __host__ __device__ void operator/= (float4 &a, float b)

inline __host__ __device__ float4 operator/ (float b, float4 a)

inline __host__ __device__ float2 fminf (float2 a, float2 b)

inline __host__ __device__ float3 fminf (float3 a, float3 b)

inline __host__ __device__ float4 fminf (float4 a, float4 b)

inline __host__ __device__ int2 min (int2 a, int2 b)

inline __host__ __device__ int3 min (int3 a, int3 b)

inline __host__ __device__ int4 min (int4 a, int4 b)

inline __host__ __device__ uint2 min (uint2 a, uint2 b)

inline __host__ __device__ uint3 min (uint3 a, uint3 b)

inline __host__ __device__ uint4 min (uint4 a, uint4 b)

inline __host__ __device__ float2 fmaxf (float2 a, float2 b)

inline __host__ __device__ float3 fmaxf (float3 a, float3 b)

inline __host__ __device__ float4 fmaxf (float4 a, float4 b)

inline __host__ __device__ int2 max (int2 a, int2 b)

inline __host__ __device__ int3 max (int3 a, int3 b)

inline __host__ __device__ int4 max (int4 a, int4 b)

inline __host__ __device__ uint2 max (uint2 a, uint2 b)

inline __host__ __device__ uint3 max (uint3 a, uint3 b)

inline __host__ __device__ uint4 max (uint4 a, uint4 b)

inline __device__ __host__ float lerp (float a, float b, float t)

inline __device__ __host__ float2 lerp (float2 a, float2 b, float t)

inline __device__ __host__ float3 lerp (float3 a, float3 b, float t)

inline __device__ __host__ float4 lerp (float4 a, float4 b, float t)

inline __device__ __host__ float clamp (float f, float a, float b)

inline __device__ __host__ int clamp (int f, int a, int b)

inline __device__ __host__ uint clamp (uint f, uint a, uint b)

inline __device__ __host__ float2 clamp (float2 v, float a, float b)

inline __device__ __host__ float2 clamp (float2 v, float2 a, float2 b)

inline __device__ __host__ float3 clamp (float3 v, float a, float b)

inline __device__ __host__ float3 clamp (float3 v, float3 a, float3 b)

inline __device__ __host__ float4 clamp (float4 v, float a, float b)

inline __device__ __host__ float4 clamp (float4 v, float4 a, float4 b)

inline __device__ __host__ int2 clamp (int2 v, int a, int b)

inline __device__ __host__ int2 clamp (int2 v, int2 a, int2 b)

inline __device__ __host__ int3 clamp (int3 v, int a, int b)

inline __device__ __host__ int3 clamp (int3 v, int3 a, int3 b)

inline __device__ __host__ int4 clamp (int4 v, int a, int b)

inline __device__ __host__ int4 clamp (int4 v, int4 a, int4 b)

inline __device__ __host__ uint2 clamp (uint2 v, uint a, uint b)

inline __device__ __host__ uint2 clamp (uint2 v, uint2 a, uint2 b)

inline __device__ __host__ uint3 clamp (uint3 v, uint a, uint b)

inline __device__ __host__ uint3 clamp (uint3 v, uint3 a, uint3 b)

inline __device__ __host__ uint4 clamp (uint4 v, uint a, uint b)

inline __device__ __host__ uint4 clamp (uint4 v, uint4 a, uint4 b)

inline __host__ __device__ float dot (float2 a, float2 b)

inline __host__ __device__ float dot (float3 a, float3 b)

inline __host__ __device__ float dot (float4 a, float4 b)

inline __host__ __device__ int dot (int2 a, int2 b)

inline __host__ __device__ int dot (int3 a, int3 b)

inline __host__ __device__ int dot (int4 a, int4 b)

inline __host__ __device__ uint dot (uint2 a, uint2 b)

inline __host__ __device__ uint dot (uint3 a, uint3 b)

inline __host__ __device__ uint dot (uint4 a, uint4 b)

inline __host__ __device__ float length (float2 v)

inline __host__ __device__ float length (float3 v)

inline __host__ __device__ float length (float4 v)

inline __host__ __device__ float2 normalize (float2 v)

inline __host__ __device__ float3 normalize (float3 v)

inline __host__ __device__ float4 normalize (float4 v)

inline __host__ __device__ float2 floorf (float2 v)

inline __host__ __device__ float3 floorf (float3 v)

inline __host__ __device__ float4 floorf (float4 v)

inline __host__ __device__ float fracf (float v)

inline __host__ __device__ float2 fracf (float2 v)

inline __host__ __device__ float3 fracf (float3 v)

inline __host__ __device__ float4 fracf (float4 v)

inline __host__ __device__ float2 fmodf (float2 a, float2 b)

inline __host__ __device__ float3 fmodf (float3 a, float3 b)

inline __host__ __device__ float4 fmodf (float4 a, float4 b)

inline __host__ __device__ float2 fabs (float2 v)

inline __host__ __device__ float3 fabs (float3 v)

inline __host__ __device__ float4 fabs (float4 v)

inline __host__ __device__ int2 abs (int2 v)

inline __host__ __device__ int3 abs (int3 v)

inline __host__ __device__ int4 abs (int4 v)

inline __host__ __device__ float3 reflect (float3 i, float3 n)

inline __host__ __device__ float3 cross (float3 a, float3 b)

inline __device__ __host__ float smoothstep (float a, float b, float x)

inline __device__ __host__ float2 smoothstep (float2 a, float2 b, float2 x)

inline __device__ __host__ float3 smoothstep (float3 a, float3 b, float3 x)

inline __device__ __host__ float4 smoothstep (float4 a, float4 b, float4 x)

file I3CLSimCUDADevice.h

#include <icetray/I3PointerTypedefs.h>

#include <string>

#include <vector>

#include <boost/shared_ptr.hpp>

Typedefs

typedef std::vector<I3CLSimCUDADevice> I3CLSimCUDADeviceSeries

Functions

bool operator==(const I3CLSimCUDADevice &a, const I3CLSimCUDADevice &b)

I3_POINTER_TYPEDEFS(I3CLSimCUDADevice)

I3_POINTER_TYPEDEFS(I3CLSimCUDADeviceSeries)

file I3CLSimHelperGenerateGeometrySource.cxx

#include “opencl/I3CLSimHelperGenerateGeometrySource.h”

#include <string>

#include <sstream>

#include <stdexcept>

#include <vector>

#include <cmath>

#include <set>

#include <icetray/I3Logging.h>

#include “dataclasses/I3Constants.h”

#include <boost/lexical_cast.hpp>

#include <boost/foreach.hpp>

#include “clsim/cl.hpp”

#include “simclasses/util/ToFloatString.h”

Defines

MAX_SUPPORTED_NUM_STRINGS

MAX_SUPPORTED_NUM_SUBDETECTORS

file I3CLSimHelperGenerateGeometrySource.h

#include <string>

#include <vector>

#include “simclasses/I3SimpleGeometry.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimHelperGenerateMediumPropertiesSource.cxx

#include <inttypes.h>

#include “opencl/I3CLSimHelperGenerateMediumPropertiesSource.h”

#include <string>

#include <sstream>

#include <stdexcept>

#include <algorithm>

#include <cmath>

#include “dataclasses/I3Constants.h”

#include <boost/lexical_cast.hpp>

#include <boost/foreach.hpp>

#include “simclasses/function/I3CLSimFunctionScatLenIceCube.h”

#include “opencl/I3CLSimHelperGenerateMediumPropertiesSource_Optimizers.h”

#include “simclasses/util/ToFloatString.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

file I3CLSimHelperGenerateMediumPropertiesSource.h

#include <string>

#include “simclasses/I3CLSimMediumProperties.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimHelperGenerateMediumPropertiesSource_Optimizers.cxx

#include “opencl/I3CLSimHelperGenerateMediumPropertiesSource_Optimizers.h”

#include <sstream>

#include <string>

#include <boost/lexical_cast.hpp>

#include <boost/foreach.hpp>

#include <boost/preprocessor.hpp>

#include “icetray/I3Units.h”

#include “simclasses/function/I3CLSimFunctionAbsLenIceCube.h”

#include “simclasses/function/I3CLSimFunctionScatLenIceCube.h”

#include “simclasses/util/ToFloatString.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

CHECK_THESE_FOR_CONSTNESS_ABSLEN

CHECK_THESE_FOR_CONSTNESS_SCATLEN

GEN_VISITOR(r, data, a)

CONST_CHECK(r, data, a)

CONST_CHECK(r, data, a)

file I3CLSimHelperGenerateMediumPropertiesSource_Optimizers.h

#include <string>

#include <vector>

#include “simclasses/function/I3CLSimFunction.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimHelperLoadProgramSource.h

#include <string>

#include <stdexcept>

#include <cstdio>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimHelperMath.cxx

#include “opencl/I3CLSimHelperMath.h”

#include “clsim/I3CLSimOpenCLDevice.h”

file I3CLSimHelperMath.h

#include <string>

file I3CLSimI3ParticleGeantConverter.hh

#include <string>

#include <cmath>

#include “G4ParticleGun.hh”

#include “G4Track.hh”

#include “G4ParticleTable.hh”

#include “G4ParticleDefinition.hh”

#include “globals.hh”

#include “G4Version.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimLightSourcePropagatorGeant4.cxx

#include “sim-services/I3LightSourceToStepConverter.h”

#include “geant4/I3CLSimLightSourcePropagatorGeant4.h”

#include <limits>

#include <deque>

#include <boost/tuple/tuple.hpp>

#include <cmath>

#include “G4RunManager.hh”

#include “G4UImanager.hh”

#include “G4StateManager.hh”

#include “G4String.hh”

#include “G4PhysListFactory.hh”

#include “G4ParticleGun.hh”

#include “G4HadronicProcessType.hh”

#include “G4Version.hh”

#include “TrkOpticalPhysics.hh”

#include “TrkDetectorConstruction.hh”

#include “TrkPrimaryGeneratorAction.hh”

#include “TrkEventAction.hh”

#include “TrkStackingAction.hh”

#include “TrkUISessionToQueue.hh”

#include “I3CLSimI3ParticleGeantConverter.hh”

#include “Randomize.hh”

#include “dataclasses/physics/I3Particle.h”

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

#include “simclasses/I3LightSource.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “simclasses/I3CLSimMediumProperties.h”

#include <stdlib.h>

#include “G4VStateDependent.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimLightSourcePropagatorGeant4.h

#include “icetray/I3Logging.h”

#include “sim-services/I3LightSourcePropagator.h”

#include “simclasses/I3CLSimQueue.h”

#include “simclasses/I3LightSource.h”

#include <map>

#include <string>

#include <atomic>

Functions

I3_POINTER_TYPEDEFS(I3CLSimLightSourcePropagatorGeant4)

file I3CLSimLightSourceToStepConverterFlasher.cxx

#include <inttypes.h>

#include <cmath>

#include <boost/lexical_cast.hpp>

#include <limits>

#include “phys-services/I3Calculator.h”

#include “clsim/I3CLSimLightSourceToStepConverterFlasher.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “clsim/I3CLSimLightSourceToStepConverterUtils.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

file I3CLSimLightSourceToStepConverterFlasher.h

#include “sim-services/I3LightSourceToStepConverter.h”

#include “simclasses/I3LightSource.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “simclasses/random_value/I3CLSimRandomValue.h”

#include “clsim/I3CLSimSpectrumTable.h”

#include <string>

#include <vector>

#include <deque>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Functions

I3_POINTER_TYPEDEFS(I3CLSimLightSourceToStepConverterFlasher)

file I3CLSimLightSourceToStepConverterPPC.cxx

#include <inttypes.h>

#include <cmath>

#include “clsim/I3CLSimLightSourceToStepConverterPPC.h”

#include “sim-services/I3SimConstants.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “clsim/I3CLSimLightSourceToStepConverterUtils.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

file I3CLSimLightSourceToStepConverterPPC.h

#include “sim-services/I3LightSourceToStepConverter.h”

#include “dataclasses/physics/I3Particle.h”

#include “simclasses/I3CLSimQueue.h”

#include <map>

#include <string>

#include <vector>

#include <deque>

#include <boost/variant.hpp>

#include <boost/thread.hpp>

#include <boost/thread/mutex.hpp>

#include <boost/thread/locks.hpp>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Functions

I3_POINTER_TYPEDEFS(I3CLSimLightSourceToStepConverterPPC)

file I3CLSimLightSourceToStepConverterUtils.cxx

#include “clsim/I3CLSimLightSourceToStepConverterUtils.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “simclasses/function/I3CLSimFunctionDeltaPeak.h”

#include “icetray/I3Units.h”

#include “dataclasses/I3Constants.h”

#include <gsl/gsl_integration.h>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

H_TIMES_C

file I3CLSimLightSourceToStepConverterUtils.h

#include “dataclasses/physics/I3Particle.h”

#include “phys-services/I3RandomService.h”

#include “dataclasses/I3Constants.h”

#include “simclasses/I3SimStep.h”

#include “simclasses/function/I3CLSimFunction.h”

#include <cmath>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimModule.cxx

#include <inttypes.h>

#include “clsim/I3CLSimModule.h”

#include <boost/foreach.hpp>

#include <boost/make_shared.hpp>

#include <boost/algorithm/string.hpp>

#include <boost/variant/get.hpp>

#include <boost/math/common_factor_rt.hpp>

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

#include “simclasses/I3CompressedPhoton.h”

#include “simclasses/function/I3CLSimFunctionConstant.h”

#include “simclasses/I3LightSource.h”

#include “clsim/I3CLSimLightSourceToStepConverterGeant4.h”

#include “clsim/I3CLSimModuleHelper.h”

#include <limits>

#include <set>

#include <deque>

#include <cmath>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3CLSimModule<I3PhotonSeriesMap>)

I3_MODULE(I3CLSimModule<I3CompressedPhotonSeriesMap>)

file I3CLSimModule.h

#include “icetray/I3Module.h”

#include “icetray/I3ConditionalModule.h”

#include “dataclasses/I3Vector.h”

#include “dataclasses/physics/I3MCTree.h”

#include “icetray/OMKey.h”

#include “dataclasses/ModuleKey.h”

#include “phys-services/I3RandomService.h”

#include “phys-services/surfaces/ExtrudedPolygon.h”

#include “clsim/I3CLSimOpenCLDevice.h”

#include “simclasses/random_value/I3CLSimRandomValue.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “simclasses/I3CLSimMediumProperties.h”

#include “clsim/I3CLSimSpectrumTable.h”

#include “simclasses/I3SimpleGeometryFromI3Geometry.h”

#include “clsim/I3CLSimStepToPhotonConverterOpenCL.h”

#include “clsim/I3CLSimLightSourceToStepConverterGeant4.h”

#include “sim-services/I3LightSourceParameterization.h”

#include “simclasses/I3Photon.h”

#include “simclasses/I3CLSimPhotonHistory.h”

#include “simclasses/I3CLSimEventStatistics.h”

#include <boost/thread.hpp>

#include <boost/thread/mutex.hpp>

#include <boost/thread/locks.hpp>

#include <vector>

#include <set>

#include <map>

#include <string>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimModuleHelper.cxx

#include <inttypes.h>

#include “clsim/I3CLSimModuleHelper.h”

#include “simclasses/function/I3CLSimFunctionConstant.h”

#include “simclasses/function/I3CLSimFunctionFromTable.h”

#include “simclasses/function/I3CLSimFunctionDeltaPeak.h”

#include “simclasses/random_value/I3CLSimRandomValueInterpolatedDistribution.h”

#include “simclasses/random_value/I3CLSimRandomValueWlenCherenkovNoDispersion.h”

#include “simclasses/random_value/I3CLSimRandomValueConstant.h”

#include <boost/foreach.hpp>

#include <boost/algorithm/string.hpp>

#include <boost/variant/get.hpp>

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

file I3CLSimModuleHelper.h

#include “phys-services/I3RandomService.h”

#include “simclasses/random_value/I3CLSimRandomValue.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “simclasses/I3CLSimMediumProperties.h”

#include “simclasses/I3SimpleGeometryFromI3Geometry.h”

#include “clsim/I3CLSimStepToPhotonConverterOpenCL.h”

#include “sim-services/I3LightSourceParameterization.h”

#include “clsim/I3CLSimOpenCLDevice.h”

#include <vector>

#include <string>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimOpenCLDevice.cxx

#include <clsim/I3CLSimOpenCLDevice.h>

#include <stdexcept>

#include <boost/foreach.hpp>

#include <boost/lexical_cast.hpp>

#include “clsim/cl.hpp”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

HAS_CL_DEVICE_FISSION

Functions

bool operator==(const I3CLSimOpenCLDevice &a, const I3CLSimOpenCLDevice &b)

file I3CLSimOpenCLDevice.h

#include “icetray/I3TrayHeaders.h”

#include <string>

#include <vector>

#include <boost/shared_ptr.hpp>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Typedefs

typedef std::vector<I3CLSimOpenCLDevice> I3CLSimOpenCLDeviceSeries

Functions

bool operator==(const I3CLSimOpenCLDevice &a, const I3CLSimOpenCLDevice &b)

I3_POINTER_TYPEDEFS(I3CLSimOpenCLDevice)

I3_POINTER_TYPEDEFS(I3CLSimOpenCLDeviceSeries)

file I3CLSimSpectrumTable.cxx

#include <clsim/I3CLSimSpectrumTable.h>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3CLSimSpectrumTable.h

#include “icetray/I3TrayHeaders.h”

#include “simclasses/function/I3CLSimFunction.h”

#include <vector>

#include <stdexcept>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Functions

I3_POINTER_TYPEDEFS(I3CLSimSpectrumTable)

file I3CLSimStepToPhotonConverterCUDA.cxx

#include “I3CLSimStepToPhotonConverterCUDA.h”

#include “simclasses/random_value/I3CLSimRandomValueInterpolatedDistribution.h”

#include “simclasses/function/I3CLSimFunctionFromTable.h”

#include “simclasses/function/I3CLSimFunctionAbsLenIceCube.h”

#include “simclasses/function/I3CLSimFunctionScatLenIceCube.h”

#include “cuda/kernels/propagationKernelSource.cuh”

#include <inttypes.h>

#include <cmath>

#include <chrono>

#include <ctime>

#include <random>

#include <icetray/I3Units.h>

#include <stdlib.h>

#include <algorithm>

#include <boost/foreach.hpp>

#include <boost/lexical_cast.hpp>

#include <boost/thread/barrier.hpp>

#include <limits>

#include <sstream>

#include <string>

#include “opencl/mwcrng_init.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

PRINTLC

DUMP_STATISTICS

file I3CLSimStepToPhotonConverterCUDA.h

#include “sim-services/I3StepToPhotonConverter.h”

#include “phys-services/I3RandomService.h”

#include <boost/thread.hpp>

#include <boost/thread/mutex.hpp>

#include <boost/thread/locks.hpp>

#include “simclasses/I3CLSimQueue.h”

#include “I3CLSimCUDADevice.h”

#include <vector>

#include <map>

#include <string>

#include <stdexcept>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

Rcs

I3CLSimStepToPhotonConverterCUDA.h 177703 2019-12-05 20:04:47Z jvansanten

Version

Rcs

177703

Date

Rcs

2019-12-05 13:04:47 -0700 (Thu, 05 Dec 2019)

Author

Claudio Kopper

Functions

I3_FORWARD_DECLARATION(I3CLSimFunctionFromTable)

I3_FORWARD_DECLARATION(I3CLSimRandomValueInterpolatedDistribution)

I3_POINTER_TYPEDEFS(I3CLSimStepToPhotonConverterCUDA)

file I3CLSimStepToPhotonConverterOpenCL.cxx

#include <inttypes.h>

#include <cmath>

#include <boost/date_time/posix_time/posix_time.hpp>

#include “clsim/I3CLSimStepToPhotonConverterOpenCL.h”

#include <chrono>

#include <string>

#include <sstream>

#include <algorithm>

#include <limits>

#include <type_traits>

#include <stdlib.h>

#include <boost/foreach.hpp>

#include <boost/lexical_cast.hpp>

#include <icetray/I3Units.h>

#include “simclasses/util/ToFloatString.h”

#include “opencl/I3CLSimHelperMath.h”

#include “opencl/I3CLSimHelperLoadProgramSource.h”

#include “opencl/I3CLSimHelperGenerateMediumPropertiesSource.h”

#include “opencl/I3CLSimHelperGenerateGeometrySource.h”

#include “opencl/mwcrng_init.h”

#include “clsim/cl.hpp”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

DUMP_STATISTICS

YIELD_TIME_MICROSECONDS

Functions

static std::string loadKernel(const std::string &name, bool header)

Variables

cl_float4 posAndTime

cl_float4 dirAndLengthAndBeta

cl_uint numPhotons

cl_float weight

cl_uint identifier

cl_uchar sourceType

cl_uchar dummy1

cl_ushort dummy2

cl_float2 dir

cl_float wavelength

cl_float cherenkovDist

cl_uint numScatters

cl_short stringID

cl_ushort omID

cl_float4 startPosAndTime

cl_float2 startDir

cl_float groupVelocity

cl_float distInAbsLens

file I3CLSimStepToPhotonConverterOpenCL.h

#include “sim-services/I3StepToPhotonConverter.h”

#include “phys-services/I3RandomService.h”

#include <boost/thread.hpp>

#include <boost/thread/mutex.hpp>

#include <boost/thread/locks.hpp>

#include “simclasses/I3CLSimQueue.h”

#include “clsim/I3CLSimOpenCLDevice.h”

#include <vector>

#include <map>

#include <string>

#include <stdexcept>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Functions

I3_POINTER_TYPEDEFS(I3CLSimStepToPhotonConverterOpenCL)

file I3CLSimStepToTableConverter.cxx

#include <fitsio.h>

#include <fitsio2.h>

#include “clsim/tabulator/I3CLSimStepToTableConverter.h”

#include “clsim/tabulator/Axes.h”

#include “simclasses/function/I3CLSimFunctionConstant.h”

#include “opencl/I3CLSimHelperMath.h”

#include “opencl/I3CLSimHelperGenerateMediumPropertiesSource.h”

#include “opencl/I3CLSimHelperLoadProgramSource.h”

#include “opencl/mwcrng_init.h”

#include “clsim/I3CLSimModuleHelper.h”

#include “clsim/I3CLSimLightSourceToStepConverterUtils.h”

#include “simclasses/util/ToFloatString.h”

#include <boost/foreach.hpp>

#include <boost/make_shared.hpp>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

Functions

I3CLSimReferenceParticle(const I3Particle &source)

Variables

uint32_t index

float weight

cl_float4 posAndTime

cl_float4 dir

cl_float4 perpDir

file I3CLSimStepToTableConverter.h

#include “simclasses/I3SimStep.h”

#include “simclasses/I3CLSimQueue.h”

#include “simclasses/I3CLSimMediumProperties.h”

#include “clsim/I3CLSimSpectrumTable.h”

#include “simclasses/random_value/I3CLSimRandomValue.h”

#include “clsim/I3CLSimOpenCLDevice.h”

#include “dataclasses/physics/I3Particle.h”

#include “clsim/tabulator/Axes.h”

#include “clsim/cl.hpp”

#include <boost/noncopyable.hpp>

#include <boost/thread.hpp>

Copyright (c) 2015 Jakob van Santen jvansanten@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$LastChangedRevision$

Date

$Date$

Author

Jakob van Santen

Functions

I3_POINTER_TYPEDEFS(I3CLSimStepToTableConverter)

file I3CLSimTabulatorModule.cxx

#include <icetray/I3Module.h>

#include <phys-services/I3RandomService.h>

#include <dataclasses/physics/I3MCTree.h>

#include “clsim/I3CLSimOpenCLDevice.h”

#include “sim-services/I3LightSourceParameterization.h”

#include “simclasses/I3CLSimMediumProperties.h”

#include “clsim/I3CLSimSpectrumTable.h”

#include “sim-services/I3LightSourceToStepConverterAsync.h”

#include “simclasses/function/I3CLSimFunction.h”

#include “clsim/I3CLSimModuleHelper.h”

#include “clsim/tabulator/I3CLSimStepToTableConverter.h”

#include “clsim/tabulator/Axes.h”

#include <boost/foreach.hpp>

#include <boost/thread.hpp>

#include <boost/make_shared.hpp>

#include <boost/scoped_ptr.hpp>

#include <boost/filesystem.hpp>

#include <fstream>

Functions

I3_MODULE(I3CLSimTabulatorModule)

file I3MCHitConverterWithIDs.cxx

#include “I3MCHitConverterWithIDs.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3MCHitConverterWithIDs.h

#include “tableio/I3Converter.h”

#include “tableio/converter/I3MapConverter.h”

#include “dataclasses/physics/I3MCHit.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Typedefs

typedef I3MapOMKeyVectorConverter<I3MCHitConverterWithIDs> I3MCHitSeriesMapConverterWithIDs

file I3MuonSlicer.cxx

#include <inttypes.h>

#include <cmath>

#include “clsim/util/I3MuonSlicer.h”

#include <boost/foreach.hpp>

#include <gsl/gsl_sys.h>

#include “dataclasses/physics/I3Particle.h”

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

#include “dataclasses/I3Constants.h”

#include “simclasses/I3MMCTrack.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3MuonSlicer)

file I3MuonSlicer.h

#include “icetray/I3Module.h”

#include “icetray/I3ConditionalModule.h”

#include <string>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3MuonSliceRemoverAndPulseRelabeler.cxx

#include <inttypes.h>

#include “icetray/I3Module.h”

#include “icetray/I3ConditionalModule.h”

#include “dataclasses/physics/I3Particle.h”

#include “dataclasses/physics/I3ParticleID.h”

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

#include “dataclasses/I3Constants.h”

#include “simclasses/I3MCPE.h”

#include “simclasses/I3Photon.h”

#include “simclasses/I3CompressedPhoton.h”

#include “simclasses/I3ParticleIDMap.hpp”

Copyright (c) 2013 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3MuonSliceRemoverAndPulseRelabeler)

file I3PhotonToMCHitConverterForMultiPMT.cxx

#include <inttypes.h>

#include “clsim/dom/I3PhotonToMCHitConverterForMultiPMT.h”

#include “simclasses/I3Photon.h”

#include “dataclasses/physics/I3MCHit.h”

#include “dataclasses/physics/I3MCTree.h”

#include “icetray/I3Logging.h”

#include “dataclasses/I3Units.h”

#include “dataclasses/I3Constants.h”

#include “dataclasses/I3Position.h”

#include “dataclasses/I3Direction.h”

#include “dataclasses/physics/I3Particle.h”

#include “dataclasses/geometry/I3Geometry.h”

#include “phys-services/I3RandomService.h”

#include <algorithm>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3PhotonToMCHitConverterForMultiPMT)

file I3PhotonToMCHitConverterForMultiPMT.h

#include <icetray/I3Module.h>

#include <icetray/I3ConditionalModule.h>

#include <icetray/I3TrayHeaders.h>

#include <icetray/I3Logging.h>

#include “phys-services/I3RandomService.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Variables

I3PhotonToMCHitConverterForMultiPMT __attribute__

file I3ShadowedPhotonRemoverModule.cxx

#include <inttypes.h>

#include <vector>

#include <iostream>

#include “clsim/shadow/I3ShadowedPhotonRemoverModule.h”

#include “simclasses/I3ExtraGeometryItemCylinder.h”

#include “simclasses/I3CompressedPhoton.h”

#include “dataclasses/I3Double.h”

#include “dataclasses/I3Constants.h”

#include “simclasses/I3CylinderMap.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3ShadowedPhotonRemoverModule)

file I3ShadowedPhotonRemoverModule.h

#include <string>

#include “icetray/I3Module.h”

#include “icetray/I3ConditionalModule.h”

#include “dataclasses/geometry/I3Geometry.h”

#include “simclasses/I3CylinderMap.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file I3TauSanitizer.cxx

#include <inttypes.h>

#include “icetray/I3Module.h”

#include “icetray/I3ConditionalModule.h”

#include <string>

#include <cmath>

#include <boost/foreach.hpp>

#include “dataclasses/physics/I3Particle.h”

#include “dataclasses/physics/I3MCTree.h”

#include “dataclasses/physics/I3MCTreeUtils.h”

#include “dataclasses/I3Constants.h”

#include “simclasses/I3MMCTrack.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

__STDC_FORMAT_MACROS

Functions

I3_MODULE(I3TauSanitizer)

file iccfeatures.h

#include <assert.h>

#include <stdint.h>

Defines

R123_GNUC_VERSION

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_NI

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_ASM_GNU

R123_USE_CPUID_MSVC

R123_USE_X86INTRIN_H

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO16_ASM

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

__STDC_CONSTANT_MACROS

file ieeehalfprecision.cxx

#include <string.h>

#include <inttypes.h>

#include “opencl/ieeehalfprecision.h”

Defines

INT16_TYPE

UINT16_TYPE

INT32_TYPE

UINT32_TYPE

Functions

int singles2halfp(void *target, const void *source, int numel)

int doubles2halfp(void *target, const void *source, int numel)

int halfp2singles(void *target, const void *source, int numel)

int halfp2doubles(void *target, const void *source, int numel)

file ieeehalfprecision.h

Functions

int singles2halfp(void *target, const void *source, int numel)

int doubles2halfp(void *target, const void *source, int numel)

int halfp2singles(void *target, const void *source, int numel)

int halfp2doubles(void *target, const void *source, int numel)

file main.cxx

#include <inttypes.h>

#include <iostream>

#include <fstream>

#include <cstddef>

#include “gmp.h”

Defines

__STDC_FORMAT_MACROS

Functions

bool is_prime(mpz_t n)

int main()

file MicroURNG.hpp

#include <stdexcept>

#include <limits>

file msvcfeatures.h

#include <assert.h>

#include <stdint.h>

Defines

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_ASM_GNU

R123_USE_CPUID_MSVC

R123_USE_X86INTRIN_H

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO16_ASM

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

__STDC_CONSTANT_MACROS

file mwcrng_init.h

#include <inttypes.h>

#include <cstdio>

#include <cstring>

#include <stdint.h>

#include <string>

#include <boost/filesystem.hpp>

#include <icetray/open.h>

#include “phys-services/I3RandomService.h”

Defines

__STDC_FORMAT_MACROS

Functions

inline int init_MWC_RNG(uint64_t *x, uint32_t *a, const uint32_t n_rng, I3RandomServicePtr randomService, std::string safeprimes_file = "")

file nvccfeatures.h

Defines

R123_USE_MULHILO64_ASM

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_GNU_UINT128

R123_ULONG_LONG

file open64features.h

#include “gccfeatures.h”

Defines

R123_USE_GNU_UINT128

R123_USE_MULHILO64_ASM

file opencl.hpp

#include <CL/opencl.h>

#include <utility>

#include <limits>

#include <iterator>

#include <mutex>

#include <cstring>

#include <functional>

#include <vector>

#include <string>

#include <memory>

#include <array>

C++ bindings for OpenCL 1.0, OpenCL 1.1, OpenCL 1.2, OpenCL 2.0, OpenCL 2.1, OpenCL 2.2, and OpenCL 3.0.

Derived from the OpenCL 1.x C++ bindings written by Benedict R. Gaster, Laurent Morichetti and Lee Howes With additions and fixes from: Brian Cole, March 3rd 2010 and April 2012 Matt Gruenke, April 2012. Bruce Merry, February 2013. Tom Deakin and Simon McIntosh-Smith, July 2013 James Price, 2015-

Optional extension support

  cl_ext_device_fission
  #define CL_HPP_USE_CL_DEVICE_FISSION
  cl_khr_d3d10_sharing
  #define CL_HPP_USE_DX_INTEROP
  cl_khr_sub_groups
  #define CL_HPP_USE_CL_SUB_GROUPS_KHR
  cl_khr_image2d_from_buffer
  #define CL_HPP_USE_CL_IMAGE2D_FROM_BUFFER_KHR

Author

Lee Howes and Bruce Merry

Version

2.2.0

Date

2019-09-18

Doxygen documentation for this header is available here:

http://khronosgroup.github.io/OpenCL-CLHPP/

The latest version of this header can be found on the GitHub releases page:

https://github.com/KhronosGroup/OpenCL-CLHPP/releases

Bugs and patches can be submitted to the GitHub repository:

https://github.com/KhronosGroup/OpenCL-CLHPP

Defines

CL_HPP_TARGET_OPENCL_VERSION

CL_TARGET_OPENCL_VERSION

CL_HPP_MINIMUM_OPENCL_VERSION

CL_USE_DEPRECATED_OPENCL_2_0_APIS

CL_USE_DEPRECATED_OPENCL_2_1_APIS

CL_USE_DEPRECATED_OPENCL_2_2_APIS

CL_HPP_NOEXCEPT_

CL_HPP_DEFINE_STATIC_MEMBER_

CL_API_PREFIX__VERSION_1_1_DEPRECATED

CL_API_SUFFIX__VERSION_1_1_DEPRECATED

CL_API_PREFIX__VERSION_1_2_DEPRECATED

CL_API_SUFFIX__VERSION_1_2_DEPRECATED

CL_CALLBACK

CL_HPP_INIT_CL_EXT_FCN_PTR_(name)

CL_HPP_INIT_CL_EXT_FCN_PTR_PLATFORM_(platform, name)

CL_HPP_ERR_STR_(x)

CL_HPP_PARAM_NAME_INFO_1_0_(F)

CL_HPP_PARAM_NAME_INFO_1_1_(F)

CL_HPP_PARAM_NAME_INFO_1_2_(F)

CL_HPP_PARAM_NAME_INFO_2_0_(F)

CL_HPP_PARAM_NAME_INFO_SUBGROUP_KHR_(F)

CL_HPP_PARAM_NAME_INFO_IL_KHR_(F)

CL_HPP_PARAM_NAME_INFO_2_1_(F)

CL_HPP_PARAM_NAME_INFO_2_2_(F)

CL_HPP_PARAM_NAME_DEVICE_FISSION_(F)

CL_HPP_PARAM_NAME_CL_KHR_EXTENDED_VERSIONING_CL3_SHARED_(F)

CL_HPP_PARAM_NAME_CL_KHR_EXTENDED_VERSIONING_KHRONLY_(F)

CL_HPP_PARAM_NAME_INFO_3_0_(F)

CL_HPP_DECLARE_PARAM_TRAITS_(token, param_name, T)

CL_HPP_PARAM_NAME_INFO_1_0_DEPRECATED_IN_2_0_(F)

CL_HPP_PARAM_NAME_INFO_1_1_DEPRECATED_IN_2_0_(F)

CL_HPP_PARAM_NAME_INFO_1_2_DEPRECATED_IN_2_0_(F)

file openclfeatures.h

Defines

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_GNU_UINT128

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

R123_USE_AES_NI

UINT64_C(x)

Typedefs

typedef ulong uint64_t

typedef uint uint32_t

typedef uchar uint8_t

file pgccfeatures.h

#include <including_a_nonexistent_file_will_stop_some_compilers_from_continuing_with_a_hopeless_task>

#include <assert.h>

#include <stdint.h>

Defines

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_ASM_GNU

R123_USE_CPUID_MSVC

R123_USE_X86INTRIN_H

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_MULHI_INTRIN

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

__STDC_CONSTANT_MACROS

file philox.h

Defines

philox2x32(c, k)

philox4x32(c, k)

Functions

_philoxNxW_tpl (2, 1, 32, uint32_t) _philoxNxW_tpl(4

file ReinterpretCtr.hpp

#include “features/compilerfeatures.h”

#include <cstring>

file sse.h

Functions

R123_STATIC_INLINE int haveAESNI ()

file sunprofeatures.h

#include <assert.h>

#include <stdint.h>

Defines

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_ASM_GNU

R123_USE_CPUID_MSVC

R123_USE_X86INTRIN_H

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO16_ASM

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

R123_USE_PHILOX_64BIT

__STDC_CONSTANT_MACROS

file threefry.h

#include “features/compilerfeatures.h”

#include “array.h”

Defines

threefry2x32(c, k)

threefry4x32(c, k)

threefry2x64(c, k)

threefry4x64(c, k)

file TrkCerenkov.cxx

#include “G4ios.hh”

#include “G4Poisson.hh”

#include “G4EmProcessSubType.hh”

#include “G4LossTableManager.hh”

#include “G4MaterialCutsCouple.hh”

#include “G4ParticleDefinition.hh”

#include “G4Event.hh”

#include “G4EventManager.hh”

#include “G4RunManager.hh”

#include “TrkUserEventInformation.hh”

#include “TrkCerenkov.hh”

#include “icetray/I3Units.h”

#include “simclasses/I3SimStep.h”

#include <boost/thread.hpp>

#include “G4Version.hh”

file TrkCerenkov.hh

#include “globals.hh”

#include “templates.hh”

#include “Randomize.hh”

#include “G4ThreeVector.hh”

#include “G4ParticleMomentum.hh”

#include “G4Step.hh”

#include “G4VProcess.hh”

#include “G4OpticalPhoton.hh”

#include “G4DynamicParticle.hh”

#include “G4Material.hh”

#include “G4PhysicsTable.hh”

#include “G4MaterialPropertiesTable.hh”

#include “G4PhysicsOrderedFreeVector.hh”

#include “simclasses/function/I3CLSimFunction.h”

file TrkDetectorConstruction.cxx

#include “TrkDetectorConstruction.hh”

#include <vector>

#include <map>

#include “G4RunManager.hh”

#include “G4LogicalBorderSurface.hh”

#include “G4LogicalSkinSurface.hh”

#include “G4MaterialTable.hh”

#include “G4Material.hh”

#include “G4Box.hh”

#include “G4UserLimits.hh”

#include “G4LogicalVolume.hh”

#include “G4ThreeVector.hh”

#include “G4PVPlacement.hh”

#include “globals.hh”

#include “G4SolidStore.hh”

#include “G4LogicalVolumeStore.hh”

#include “G4PhysicalVolumeStore.hh”

#include “G4GeometryManager.hh”

#include “G4NistManager.hh”

#include “G4SystemOfUnits.hh”

#include “G4PhysicalConstants.hh”

#include “icetray/I3Logging.h”

#include “icetray/I3Units.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkDetectorConstruction.hh

#include “G4Material.hh”

#include “G4RotationMatrix.hh”

#include “G4VUserDetectorConstruction.hh”

#include “simclasses/I3CLSimMediumProperties.h”

#include <map>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

TrkDetectorConstruction_H

file TrkEventAction.cxx

#include “TrkEventAction.hh”

#include “TrkUserEventInformation.hh”

#include “G4EventManager.hh”

#include “G4SDManager.hh”

#include “G4RunManager.hh”

#include “G4Event.hh”

#include “G4TrajectoryContainer.hh”

#include “G4Trajectory.hh”

#include “G4VVisManager.hh”

#include “G4ios.hh”

#include “G4UImanager.hh”

#include “G4UnitsTable.hh”

#include “G4ParticleTable.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkEventAction.hh

#include “G4UserEventAction.hh”

#include “globals.hh”

#include “geant4/I3CLSimLightSourcePropagatorGeant4.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkOpticalPhysics.cxx

#include “TrkOpticalPhysics.hh”

#include “globals.hh”

#include “G4ios.hh”

#include <iomanip>

#include “G4ProcessManager.hh”

#include “TrkCerenkov.hh”

#include “G4ParticleDefinition.hh”

#include “G4ParticleTable.hh”

#include “G4OpticalPhoton.hh”

#include “G4Version.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkOpticalPhysics.hh

#include “G4VPhysicsConstructor.hh”

#include “simclasses/function/I3CLSimFunction.h”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkPrimaryGeneratorAction.cxx

#include “TrkPrimaryGeneratorAction.hh”

#include “G4Event.hh”

#include “G4ParticleGun.hh”

#include “globals.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkPrimaryGeneratorAction.hh

#include “G4GeneralParticleSource.hh”

#include “G4VUserPrimaryGeneratorAction.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

TrkPrimaryGeneratorAction_h

file TrkStackingAction.cxx

#include “TrkStackingAction.hh”

#include “TrkUserEventInformation.hh”

#include “G4ios.hh”

#include “G4ParticleDefinition.hh”

#include “G4ParticleTypes.hh”

#include “G4Track.hh”

#include “G4Material.hh”

#include “G4MaterialPropertyVector.hh”

#include “G4RunManager.hh”

#include “G4Event.hh”

#include “G4EventManager.hh”

#include “G4UnitsTable.hh”

#include “I3CLSimI3ParticleGeantConverter.hh”

#include “G4Version.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkStackingAction.hh

#include “globals.hh”

#include “G4UserStackingAction.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkUISessionToQueue.cxx

#include “TrkUISessionToQueue.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkUISessionToQueue.hh

#include “globals.hh”

#include “G4UIsession.hh”

#include “G4Version.hh”

#include <boost/shared_ptr.hpp>

#include “simclasses/I3CLSimQueue.h”

#include <string>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

Defines

TrkUISessionToQueue_H

file TrkUserEventInformation.cxx

#include “TrkUserEventInformation.hh”

#include “TrkDetectorConstruction.hh”

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file TrkUserEventInformation.hh

#include “G4VUserEventInformation.hh”

#include “globals.hh”

#include “geant4/I3CLSimLightSourcePropagatorGeant4.h”

#include “simclasses/I3LightSource.h”

#include <deque>

#include <boost/tuple/tuple.hpp>

#include <boost/thread.hpp>

#include <sys/time.h>

#include <sys/resource.h>

#include “icetray/I3Units.h”

#include <map>

Copyright (c) 2011, 2012 Claudio Kopper claudio.kopper@icecube.wisc.edu and the IceCube Collaboration http://www.icecube.wisc.edu

Permission to use, copy, modify, and/or distribute this software for any purpose with or without fee is hereby granted, provided that the above copyright notice and this permission notice appear in all copies.

THE SOFTWARE IS PROVIDED “AS IS” AND THE AUTHOR DISCLAIMS ALL WARRANTIES WITH REGARD TO THIS SOFTWARE INCLUDING ALL IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS. IN NO EVENT SHALL THE AUTHOR BE LIABLE FOR ANY SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.

$Id$

Version

$Revision$

Date

$Date$

Author

Claudio Kopper

file xlcfeatures.h

#include <including_a_nonexistent_file_will_stop_some_compilers_from_continuing_with_a_hopeless_task>

#include <assert.h>

#include <stdint.h>

Defines

R123_STATIC_INLINE

R123_FORCE_INLINE(decl)

R123_CUDA_DEVICE

R123_ASSERT(x)

R123_BUILTIN_EXPECT(expr, likely)

R123_USE_AES_NI

R123_USE_SSE4_2

R123_USE_SSE4_1

R123_USE_SSE

R123_USE_AES_OPENSSL

R123_USE_GNU_UINT128

R123_USE_ASM_GNU

R123_USE_CPUID_MSVC

R123_USE_X86INTRIN_H

R123_USE_IA32INTRIN_H

R123_USE_XMMINTRIN_H

R123_USE_EMMINTRIN_H

R123_USE_SMMINTRIN_H

R123_USE_WMMINTRIN_H

R123_USE_INTRIN_H

R123_USE_MULHILO32_ASM

R123_USE_MULHILO64_MULHI_INTRIN

R123_MULHILO64_MULHI_INTRIN

R123_USE_MULHILO32_MULHI_INTRIN

R123_MULHILO32_MULHI_INTRIN

R123_USE_MULHILO64_ASM

R123_USE_MULHILO64_MSVC_INTRIN

R123_USE_MULHILO64_CUDA_INTRIN

R123_USE_MULHILO64_OPENCL_INTRIN

__STDC_CONSTANT_MACROS

page porting

The Random123 library is portable across C, C++, CUDA, OpenCL environments, and multiple operating systems (Linux, Windows 7, Mac OS X, FreeBSD, Solaris). This level of portability requires the abstraction of some features and idioms that are either not standardized (e.g., asm statments), or for which different vendors have their own standards (e.g., SSE intrinsics) or for which vendors simply refuse to conform to well-established standards (e.g., <inttypes.h>).

Random123/features/compilerfeatures.h conditionally includes a compiler-or-OS-specific Random123/featires/XXXfeatures.h file which defines appropriate values for the preprocessor symbols which can be used with a specific compiler or OS. Those symbols will then be used by other header files and source files in the Random123 library (and may be used by applications) to control what actually gets presented to the compiler.

Most of the symbols are boolean valued. In general, they will always be defined with value either 1 or 0, so do NOT use #ifdef. Use #if R123_USE_SOMETHING instead.

Library users can override any value by defining the pp-symbol with a compiler option, e.g.,

cc -DR123_USE_MULHILO64_C99

will use a strictly c99 version of the full-width 64x64->128-bit multiplication function, even if it would be disabled by default.

All boolean-valued pre-processor symbols in Random123/features/compilerfeatures.h start with the prefix R123_USE_

         AES_NI
         AES_OPENSSL
         SSE4_2
         SSE4_1
         SSE

         STD_RANDOM

         GNU_UINT128
         ASM_GNU
         ASM_MSASM

         CPUID_MSVC

         CXX11_RANDOM
         CXX11_TYPE_TRAITS
         CXX11_STATIC_ASSERT
         CXX11_CONSTEXPR
         CXX11_UNRESTRICTED_UNIONS
         CXX11_EXPLICIT_CONVERSIONS
         CXX11_LONG_LONG
         CXX11 
   
         X86INTRIN_H
         IA32INTRIN_H
         XMMINTRIN_H
         EMMINTRIN_H
         SMMINTRIN_H
         WMMINTRIN_H
         INTRIN_H

         MULHILO32_ASM
         MULHILO64_ASM
         MULHILO64_MSVC_INTRIN
         MULHILO64_CUDA_INTRIN
         MULHILO64_OPENCL_INTRIN
         MULHILO64_C99

         U01_DOUBLE

Most have obvious meanings. Some non-obvious ones:

AES_NI and AES_OPENSSL are not mutually exclusive. You can have one, both or neither.

GNU_UINT128 says that it’s safe to use __uint128_t, but it does not require its use. In particular, it should be used in mulhilo<uint64_t> only if MULHILO64_ASM is unset.

If the XXXINTRIN_H macros are true, then one should

#include <xxxintrin.h>

to gain accesss to compiler intrinsics.

The CXX11_SOME_FEATURE macros allow the code to use specific features of the C++11 language and library. The catchall In the absence of a specific CXX11_SOME_FEATURE, the feature is controlled by the catch-all R123_USE_CXX11 macro.

U01_DOUBLE defaults on, and can be turned off (set to 0) if one does not want the utility functions that convert to double (i.e. u01_*_53()), e.g. on OpenCL without the cl_khr_fp64 extension.

There are a number of invariants that are always true. Application code may choose to rely on these:

• ASM_GNU and ASM_MASM are mutually exclusive

• The “higher” SSE values imply the lower ones.

There are also non-boolean valued symbols:

• R123_STATIC_INLINE - According to both C99 and GNU99, the ‘static inline’ declaration allows the compiler to not emit code if the function is not used. Note that the semantics of ‘inline’, ‘static’ and ‘extern’ in gcc have changed over time and are subject to modification by command line options, e.g., -std=gnu89, -fgnu-inline. Nevertheless, it appears that the meaning of ‘static inline’ has not changed over time and (with a little luck) the use of ‘static inline’ here will be portable between versions of gcc and to other C99 compilers. See: http://gcc.gnu.org/onlinedocs/gcc/Inline.html http://www.greenend.org.uk/rjk/2003/03/inline.html

• R123_FORCE_INLINE(decl) - which expands to ‘decl’, adorned with the compiler-specific embellishments to strongly encourage that the declared function be inlined. If there is no such compiler-specific magic, it should expand to decl, unadorned.

• R123_CUDA_DEVICE - which expands to device (or something else with sufficiently similar semantics) when CUDA is in use, and expands to nothing in other cases.

• R123_ASSERT(x) - which expands to assert(x), or maybe to nothing at all if we’re in an environment so feature-poor that you can’t even call assert (I’m looking at you, CUDA and OpenCL), or even include assert.h safely (OpenCL).

• R123_STATIC_ASSERT(expr,msg) - which expands to static_assert(expr,msg), or to an expression that will raise a compile-time exception if expr is not true.

• R123_ULONG_LONG - which expands to a declaration of the longest available unsigned integer.

• R123_64BIT(x) - expands to something equivalent to UINT64_C(x) from <stdint.h>, even in environments where <stdint.h> is not available, e.g., MSVC and OpenCL.

• R123_BUILTIN_EXPECT(expr,likely_value) - expands to something with the semantics of gcc’s __builtin_expect(expr,likely_value). If the environment has nothing like __builtin_expect, it should expand to just expr.

dir clsim

dir clsim

dir clsim

dir conventional

dir converter

dir cuda

dir dom

dir dom

dir features

dir geant4

dir icetray

dir include

dir kernels

dir kernels

dir make_safeprimes

dir opencl

dir private

dir public

dir Random123

dir Random123

dir resources

dir shadow

dir shadow

dir tabulator

dir util

dir util

page index

Introduction

For many large applications C++ is the language of choice and so it seems reasonable to define C++ bindings for OpenCL.

The interface is contained with a single C++ header file opencl.hpp and all definitions are contained within the namespace cl. There is no additional requirement to include cl.h and to use either the C++ or original C bindings; it is enough to simply include opencl.hpp.

The bindings themselves are lightweight and correspond closely to the underlying C API. Using the C++ bindings introduces no additional execution overhead.

There are numerous compatibility, portability and memory management fixes in the new header as well as additional OpenCL 2.0 features. As a result the header is not directly backward compatible and for this reason we release it as opencl.hpp rather than a new version of cl.hpp.

Compatibility

Due to the evolution of the underlying OpenCL API the 2.0 C++ bindings include an updated approach to defining supported feature versions and the range of valid underlying OpenCL runtime versions supported.

The combination of preprocessor macros CL_HPP_TARGET_OPENCL_VERSION and CL_HPP_MINIMUM_OPENCL_VERSION control this range. These are three digit decimal values representing OpenCL runime versions. The default for the target is 200, representing OpenCL 2.0 and the minimum is also defined as 200. These settings would use 2.0 API calls only. If backward compatibility with a 1.2 runtime is required, the minimum version may be set to 120.

Note that this is a compile-time setting, and so affects linking against a particular SDK version rather than the versioning of the loaded runtime.

The earlier versions of the header included basic vector and string classes based loosely on STL versions. These were difficult to maintain and very rarely used. For the 2.0 header we now assume the presence of the standard library unless requested otherwise. We use std::array, std::vector, std::shared_ptr and std::string throughout to safely manage memory and reduce the chance of a recurrance of earlier memory management bugs.

These classes are used through typedefs in the cl namespace: cl::array, cl::vector, cl::pointer and cl::string. In addition cl::allocate_pointer forwards to std::allocate_shared by default. In all cases these standard library classes can be replaced with custom interface-compatible versions using the CL_HPP_NO_STD_ARRAY, CL_HPP_NO_STD_VECTOR, CL_HPP_NO_STD_UNIQUE_PTR and CL_HPP_NO_STD_STRING macros.

The OpenCL 1.x versions of the C++ bindings included a size_t wrapper class to interface with kernel enqueue. This caused unpleasant interactions with the standard size_t declaration and led to namespacing bugs. In the 2.0 version we have replaced this with a std::array-based interface. However, the old behaviour can be regained for backward compatibility using the CL_HPP_ENABLE_SIZE_T_COMPATIBILITY macro.

Finally, the program construction interface used a clumsy vector-of-pairs design in the earlier versions. We have replaced that with a cleaner vector-of-vectors and vector-of-strings design. However, for backward compatibility old behaviour can be regained with the CL_HPP_ENABLE_PROGRAM_CONSTRUCTION_FROM_ARRAY_COMPATIBILITY macro.

In OpenCL 2.0 OpenCL C is not entirely backward compatibility with earlier versions. As a result a flag must be passed to the OpenCL C compiled to request OpenCL 2.0 compilation of kernels with 1.2 as the default in the absence of the flag. In some cases the C++ bindings automatically compile code for ease. For those cases the compilation defaults to OpenCL C 2.0. If this is not wanted, the CL_HPP_CL_1_2_DEFAULT_BUILD macro may be specified to assume 1.2 compilation. If more fine-grained decisions on a per-kernel bases are required then explicit build operations that take the flag should be used.

Parameters

This header may be parameterized by a set of preprocessor macros.

• CL_HPP_TARGET_OPENCL_VERSION

  Defines the target OpenCL runtime version to build the header against. Defaults to 200, representing OpenCL 2.0.

• CL_HPP_NO_STD_STRING

  Do not use the standard library string class. cl::string is not defined and may be defined by the user before opencl.hpp is included.

• CL_HPP_NO_STD_VECTOR

  Do not use the standard library vector class. cl::vector is not defined and may be defined by the user before opencl.hpp is included.

• CL_HPP_NO_STD_ARRAY

  Do not use the standard library array class. cl::array is not defined and may be defined by the user before opencl.hpp is included.

• CL_HPP_NO_STD_UNIQUE_PTR

  Do not use the standard library unique_ptr class. cl::pointer and the cl::allocate_pointer functions are not defined and may be defined by the user before opencl.hpp is included.

• CL_HPP_ENABLE_DEVICE_FISSION

  Enables device fission for OpenCL 1.2 platforms.

• CL_HPP_ENABLE_EXCEPTIONS

  Enable exceptions for use in the C++ bindings header. This is the preferred error handling mechanism but is not required.

• CL_HPP_ENABLE_SIZE_T_COMPATIBILITY

  Backward compatibility option to support cl.hpp-style size_t class. Replaces the updated std::array derived version and removal of size_t from the namespace. Note that in this case the new size_t class is placed in the cl::compatibility namespace and thus requires an additional using declaration for direct backward compatibility.

• CL_HPP_ENABLE_PROGRAM_CONSTRUCTION_FROM_ARRAY_COMPATIBILITY

  Enable older vector of pairs interface for construction of programs.

• CL_HPP_CL_1_2_DEFAULT_BUILD

  Default to OpenCL C 1.2 compilation rather than OpenCL C 2.0 applies to use of cl::Program construction and other program build variants.

• CL_HPP_USE_CL_SUB_GROUPS_KHR

  Enable the cl_khr_subgroups extension.

• CL_HPP_USE_IL_KHR

  Enable the cl_khr_il_program extension.

Example

The following example shows a general use case for the C++ bindings, including support for the optional exception feature and also the supplied vector and string classes, see following sections for decriptions of these features.

#define CL_HPP_ENABLE_EXCEPTIONS
#define CL_HPP_TARGET_OPENCL_VERSION 200

#include <CL/opencl.hpp>
#include <iostream>
#include <vector>
#include <memory>
#include <algorithm>

const int numElements = 32;

int main(void)
{
    // Filter for a 2.0 platform and set it as the default
    std::vector<cl::Platform> platforms;
    cl::Platform::get(&platforms);
    cl::Platform plat;
    for (auto &p : platforms) {
        std::string platver = p.getInfo<CL_PLATFORM_VERSION>();
        if (platver.find("OpenCL 2.") != std::string::npos) {
            plat = p;
        }
    }
    if (plat() == 0)  {
        std::cout << "No OpenCL 2.0 platform found.";
        return -1;
    }

    cl::Platform newP = cl::Platform::setDefault(plat);
    if (newP != plat) {
        std::cout << "Error setting default platform.";
        return -1;
    }

    // Use C++11 raw string literals for kernel source code
    std::string kernel1{R"CLC(
        global int globalA;
        kernel void updateGlobal()
        {
          globalA = 75;
        }
    )CLC"};
    std::string kernel2{R"CLC(
        typedef struct { global int *bar; } Foo;
        kernel void vectorAdd(global const Foo* aNum, global const int *inputA, global const int *inputB,
                              global int *output, int val, write_only pipe int outPipe, queue_t childQueue)
        {
          output[get_global_id(0)] = inputA[get_global_id(0)] + inputB[get_global_id(0)] + val + *(aNum->bar);
          write_pipe(outPipe, &val);
          queue_t default_queue = get_default_queue();
          ndrange_t ndrange = ndrange_1D(get_global_size(0)/2, get_global_size(0)/2);

          // Have a child kernel write into third quarter of output
          enqueue_kernel(default_queue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange,
            ^{
                output[get_global_size(0)*2 + get_global_id(0)] =
                  inputA[get_global_size(0)*2 + get_global_id(0)] + inputB[get_global_size(0)*2 + get_global_id(0)] + globalA;
            });

          // Have a child kernel write into last quarter of output
          enqueue_kernel(childQueue, CLK_ENQUEUE_FLAGS_WAIT_KERNEL, ndrange,
            ^{
                output[get_global_size(0)*3 + get_global_id(0)] =
                  inputA[get_global_size(0)*3 + get_global_id(0)] + inputB[get_global_size(0)*3 + get_global_id(0)] + globalA + 2;
            });
        }
    )CLC"};

    // New simpler string interface style
    std::vector<std::string> programStrings {kernel1, kernel2};

    cl::Program vectorAddProgram(programStrings);
    try {
        vectorAddProgram.build("-cl-std=CL2.0");
    }
    catch (...) {
        // Print build info for all devices
        cl_int buildErr = CL_SUCCESS;
        auto buildInfo = vectorAddProgram.getBuildInfo<CL_PROGRAM_BUILD_LOG>(&buildErr);
        for (auto &pair : buildInfo) {
            std::cerr << pair.second << std::endl << std::endl;
        }

        return 1;
    }

    typedef struct { int *bar; } Foo;

    // Get and run kernel that initializes the program-scope global
    // A test for kernels that take no arguments
    auto program2Kernel =
        cl::KernelFunctor<>(vectorAddProgram, "updateGlobal");
    program2Kernel(
        cl::EnqueueArgs(
        cl::NDRange(1)));

    // SVM allocations

    auto anSVMInt = cl::allocate_svm<int, cl::SVMTraitCoarse<>>();
    *anSVMInt = 5;
    cl::SVMAllocator<Foo, cl::SVMTraitCoarse<cl::SVMTraitReadOnly<>>> svmAllocReadOnly;
    auto fooPointer = cl::allocate_pointer<Foo>(svmAllocReadOnly);
    fooPointer->bar = anSVMInt.get();
    cl::SVMAllocator<int, cl::SVMTraitCoarse<>> svmAlloc;
    std::vector<int, cl::SVMAllocator<int, cl::SVMTraitCoarse<>>> inputA(numElements, 1, svmAlloc);
    cl::coarse_svm_vector<int> inputB(numElements, 2, svmAlloc);

    //

    // Traditional cl_mem allocations
    std::vector<int> output(numElements, 0xdeadbeef);
    cl::Buffer outputBuffer(begin(output), end(output), false);
    cl::Pipe aPipe(sizeof(cl_int), numElements / 2);

    // Default command queue, also passed in as a parameter
    cl::DeviceCommandQueue defaultDeviceQueue = cl::DeviceCommandQueue::makeDefault(
        cl::Context::getDefault(), cl::Device::getDefault());

    auto vectorAddKernel =
        cl::KernelFunctor<
            decltype(fooPointer)&,
            int*,
            cl::coarse_svm_vector<int>&,
            cl::Buffer,
            int,
            cl::Pipe&,
            cl::DeviceCommandQueue
            >(vectorAddProgram, "vectorAdd");

    // Ensure that the additional SVM pointer is available to the kernel
    // This one was not passed as a parameter
    vectorAddKernel.setSVMPointers(anSVMInt);

    // Hand control of coarse allocations to runtime
    cl::enqueueUnmapSVM(anSVMInt);
    cl::enqueueUnmapSVM(fooPointer);
    cl::unmapSVM(inputB);
    cl::unmapSVM(output2);

        cl_int error;
        vectorAddKernel(
        cl::EnqueueArgs(
            cl::NDRange(numElements/2),
            cl::NDRange(numElements/2)),
        fooPointer,
        inputA.data(),
        inputB,
        outputBuffer,
        3,
        aPipe,
        defaultDeviceQueue,
           error
        );

    cl::copy(outputBuffer, begin(output), end(output));
    // Grab the SVM output vector using a map
    cl::mapSVM(output2);

    cl::Device d = cl::Device::getDefault();

    std::cout << "Output:\n";
    for (int i = 1; i < numElements; ++i) {
        std::cout << "\t" << output[i] << "\n";
    }
    std::cout << "\n\n";

    return 0;
}