Getting to know how Chroma uses PyCUDA
=========================================


* http://mathema.tician.de/software/pycuda
* https://developer.nvidia.com/category/zone/cuda-zone


PyCUDA compilation in Chroma
-------------------------------

CUDA source is compiled via `pycuda.compiler.SourceModule`::

    simon:chroma blyth$ find . -name '*.py' -exec grep -H SourceModule {} \;
    ./chroma/gpu/tools.py:    """Returns a pycuda.compiler.SourceModule object from a CUDA source file
    ./chroma/gpu/tools.py:    return pycuda.compiler.SourceModule(source, options=options,
    ./chroma/gpu/tools.py:    module = pycuda.compiler.SourceModule(init_rng_src, no_extern_c=True)
    ./test/linalg_test.py:from pycuda.compiler import SourceModule
    ./test/linalg_test.py:mod = SourceModule(source, options=['-I' + source_directory], no_extern_c=True, cache_dir=False)
    ./test/matrix_test.py:from pycuda.compiler import SourceModule
    ./test/matrix_test.py:mod = SourceModule(source, options=['-I' + source_directory], no_extern_c=True, cache_dir=False)
    ./test/rotate_test.py:from pycuda.compiler import SourceModule
    ./test/rotate_test.py:mod = SourceModule(source, options=['-I' + source_directory], no_extern_c=True)
    ./test/test_sample_cdf.py:from pycuda.compiler import SourceModule
    ./test/test_sample_cdf.py:        self.mod = SourceModule(source, options=['-I' + source_directory], no_extern_c=True, cache_dir=False)
    simon:chroma blyth$ 



CUDA modules available to python
-----------------------------------

::

    simon:chroma blyth$ find . -name '*.py' -exec grep -H get_cu_module {} \;
    ./chroma/benchmark.py:    module = gpu.get_cu_module('mesh.h', options=('--use_fast_math',))

    ./chroma/camera.py:        self.gpu_funcs = gpu.GPUFuncs(gpu.get_cu_module('mesh.h'))
    ./chroma/camera.py:        self.hybrid_funcs = gpu.GPUFuncs(gpu.get_cu_module('hybrid_render.cu'))

    ./chroma/gpu/bvh.py:from chroma.gpu.tools import get_cu_module, cuda_options, \
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,
    ./chroma/gpu/bvh.py:    bvh_module = get_cu_module('bvh.cu', options=cuda_options,

    ./chroma/gpu/daq.py:from chroma.gpu.tools import get_cu_module, cuda_options, GPUFuncs, \
    ./chroma/gpu/daq.py:        self.module = get_cu_module('daq.cu', options=cuda_options, 

    ./chroma/gpu/detector.py:from chroma.gpu.tools import get_cu_module, get_cu_source, cuda_options, \

    ./chroma/gpu/geometry.py:from chroma.gpu.tools import get_cu_module, get_cu_source, cuda_options, \
    ./chroma/gpu/geometry.py:        module = get_cu_module('mesh.h', options=cuda_options)

    ./chroma/gpu/pdf.py:from chroma.gpu.tools import get_cu_module, cuda_options, GPUFuncs, chunk_iterator
    ./chroma/gpu/pdf.py:        self.module = get_cu_module('pdf.cu', options=cuda_options,
    ./chroma/gpu/pdf.py:        self.module = get_cu_module('pdf.cu', options=cuda_options,

    ./chroma/gpu/photon.py:from chroma.gpu.tools import get_cu_module, cuda_options, GPUFuncs, \
    ./chroma/gpu/photon.py:        module = get_cu_module('propagate.cu', options=cuda_options)
    ./chroma/gpu/photon.py:        module = get_cu_module('propagate.cu', options=cuda_options)

    ./chroma/gpu/render.py:from chroma.gpu.tools import get_cu_module, cuda_options, GPUFuncs, \
    ./chroma/gpu/render.py:        transform_module = get_cu_module('transform.cu', options=cuda_options)
    ./chroma/gpu/render.py:        render_module = get_cu_module('render.cu', options=cuda_options)

    ./chroma/gpu/tools.py:def get_cu_module(name, options=None, include_source_directory=True):

    ./test/test_ray_intersection.py:        self.module = chroma.gpu.get_cu_module('mesh.h')




::

    simon:chroma blyth$ find . -name '*.cu'
    ./chroma/cuda/bvh.cu
    ./chroma/cuda/daq.cu
    ./chroma/cuda/hybrid_render.cu
    ./chroma/cuda/pdf.cu
    ./chroma/cuda/propagate.cu
    ./chroma/cuda/render.cu
    ./chroma/cuda/tools.cu
    ./chroma/cuda/transform.cu

    ./test/linalg_test.cu
    ./test/matrix_test.cu
    ./test/rotate_test.cu
    ./test/test_sample_cdf.cu
    simon:chroma blyth$