Hi, I just started a few days ago on a similar problem (though linear
interpolation is fine for me). See the below code snippet to get a
feel for what's going on with textures, though I strongly recommend
reading the 4-5 sections on the subject scattered throughout the CUDA
Programming Guide.
This example gets more complicated if like me you want to use
complex-valued data, but I'm guessing you don't. It just demonstrates
copying a Numpy array to a texture and some linear interpolation with
normalized indexing in the kernel that writes to a standard
non-textured array in global memory.
Note that set_filter_mode() isn't mentioned in the docs.
### Start. Sorry, I just today learned about PEP-8 :(
import pycuda.driver as cuda
import pycuda.autoinit
import numpy
realrow = numpy.array([1.0, 2.0, 3.0, 4.0, 5.0],
dtype=numpy.float32).reshape(1,5)
mod_copy_texture=cuda.SourceModule("""
texture<float, 1> tex;
__global__ void copy_texture_kernel(float * data) {
int ty=threadIdx.y;
data[ty] = tex1D(tex, (float)(ty)/2.0f);
}
""")
copy_texture_func =
mod_copy_texture.get_function("copy_texture_kernel")
texref = mod_copy_texture.get_texref("tex")
cuda.matrix_to_texref(realrow, texref, order="C")
texref.set_flags(cuda.TRSF_NORMALIZED_COORDINATES)
texref.set_filter_mode(cuda.filter_mode.LINEAR)
gpu_output = numpy.zeros_like(realrow)
copy_texture_func(cuda.Out(gpu_output), block=(1,5,1),
texrefs=[texref])
print "Output:"
print gpu_output
### End
Thanks to Andreas for sorting me out so far :), I appreciate it.
Ahmed
On Tue, Mar 10, 2009 at 12:39 PM, Holger Rapp <[email protected]>
wrote:
Hey,
I discovered pycuda today (after deciding it is time to finally
put my
cuda card to some use) and was very thrilled to find a library so
much in
my style: write code in a nice language, make sure the inner loop is
fast! I was also quite pleased how fast I got my toy example (1d
cubic
spline interpolation) to run on the gpu. I then turned to my real
problem.
I need cubic interpolation of numpy arrays, so I can sample my
pixels at
pixel positions (x=2.345,y=pi). I used scipy.interpolate before,
but now
I'm looking for a cuda implementation. I found
http://www.dannyruijters.nl/cubicinterpolation/ which seems like
exactly
what i want, but i was unable to get it to work with pycuda: The
kernels
rely on texture<> to write their results to and I have not
understood how
I can feed the memory of my numpy array as a texture to the
kernels. I
have some vague understanding of textures in cuda in general, so I
think
some preprocessing is needed (feeding alignement informations, how
the
data should be adressed and it must be transferred to the cuda
device).
Does somebody have sample code using pycuda? Something simple like a
rotation kernel would be a perfect example!
Hope somebody can help!
Greetings and thanks for the great work with pycuda. I will follow
it
closely.
Holger
----------------------------------------
Dipl.-Phys. Holger Rapp
Institut für Mess- und Regelungstechnik
Universität Karlsruhe (TH)
Engler-Bunte-Ring 21
76131 Karlsruhe, Germany
Geb. 40.32, Zi. 232, zweite Etage
Tel: +49 (0)721 / 608-2341
Fax: +49 (0)721 / 661874
Mail: [email protected]
Web: www.mrt.uni-karlsruhe.de
----------------------------------------
_______________________________________________
PyCuda mailing list
[email protected]
http://tiker.net/mailman/listinfo/pycuda_tiker.net
_______________________________________________
PyCuda mailing list
[email protected]
http://tiker.net/mailman/listinfo/pycuda_tiker.net