From 142b3c3caff164deb9bc7b2848e58e52387723ff Mon Sep 17 00:00:00 2001
From: Stan Seibert <stan@mtrr.org>
Date: Fri, 16 Sep 2011 15:02:02 -0400
Subject: Move CUDA source inside chroma package, rename tests directory to
 test

---
 test/rotate_test.py | 67 +++++++++++++++++++++++++++++++++++++++++++++++++++++
 1 file changed, 67 insertions(+)
 create mode 100644 test/rotate_test.py

(limited to 'test/rotate_test.py')

diff --git a/test/rotate_test.py b/test/rotate_test.py
new file mode 100644
index 0000000..92eff84
--- /dev/null
+++ b/test/rotate_test.py
@@ -0,0 +1,67 @@
+import os
+import numpy as np
+from pycuda import autoinit
+from pycuda.compiler import SourceModule
+import pycuda.driver as cuda
+from pycuda import gpuarray
+float3 = gpuarray.vec.float3
+
+def rotate(x, phi, n):
+    x = np.asarray(x)
+    n = np.asarray(n)
+
+    r = np.cos(phi)*np.identity(3) + (1-np.cos(phi))*np.outer(n,n) + \
+        np.sin(phi)*np.array([[0,n[2],-n[1]],[-n[2],0,n[0]],[n[1],-n[0],0]])
+
+    return np.inner(x,r)
+
+print 'device %s' % autoinit.device.name()
+
+current_directory = os.path.split(os.path.realpath(__file__))[0]
+source_directory = current_directory + '/../src'
+
+source = open(current_directory + '/rotate_test.cu').read()
+
+mod = SourceModule(source, options=['-I' + source_directory], no_extern_c=True, cache_dir=False)
+
+rotate_gpu = mod.get_function('rotate')
+
+size = {'block': (100,1,1), 'grid': (1,1)}
+
+a = np.empty(size['block'][0], dtype=float3)
+n = np.empty(size['block'][0], dtype=float3)
+phi = np.random.random_sample(size=a.size).astype(np.float32)
+
+a['x'] = np.random.random_sample(size=a.size)
+a['y'] = np.random.random_sample(size=a.size)
+a['z'] = np.random.random_sample(size=a.size)
+
+n['x'] = np.random.random_sample(size=n.size)
+n['y'] = np.random.random_sample(size=n.size)
+n['z'] = np.random.random_sample(size=n.size)
+
+a['x'] = np.ones(a.size)
+a['y'] = np.zeros(a.size)
+a['z'] = np.zeros(a.size)
+
+n['x'] = np.zeros(n.size)
+n['y'] = np.zeros(n.size)
+n['z'] = np.ones(n.size)
+
+phi = np.array([np.pi/2]*a.size).astype(np.float32)
+
+def testrotate():
+    dest = np.empty(a.size, dtype=float3)
+    rotate_gpu(cuda.In(a), cuda.In(phi), cuda.In(n), cuda.Out(dest), **size)
+    for v, theta, w, rdest in zip(a,phi,n,dest):
+        r = rotate((v['x'], v['y'], v['z']), theta, (w['x'], w['y'], w['z']))
+        if not np.allclose(rdest['x'], r[0]) or \
+                not np.allclose(rdest['y'], r[1]) or \
+                not np.allclose(rdest['z'], r[2]):
+            print v
+            print theta
+            print w
+            print r
+            print rdest
+            assert False
+
-- 
cgit 


From a2f7a3aed3eccb018dda989b5390371d6c0d6784 Mon Sep 17 00:00:00 2001
From: Stan Seibert <stan@mtrr.org>
Date: Fri, 16 Sep 2011 19:21:57 -0400
Subject: Add a srcdir attribute to the chroma.cuda module.  This consolidates
 the calculation of the path to the CUDA source code in one place.

---
 test/rotate_test.py | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

(limited to 'test/rotate_test.py')

diff --git a/test/rotate_test.py b/test/rotate_test.py
index 92eff84..7ac7804 100644
--- a/test/rotate_test.py
+++ b/test/rotate_test.py
@@ -18,7 +18,7 @@ def rotate(x, phi, n):
 print 'device %s' % autoinit.device.name()
 
 current_directory = os.path.split(os.path.realpath(__file__))[0]
-source_directory = current_directory + '/../src'
+from chroma.cuda import srcdir as source_directory
 
 source = open(current_directory + '/rotate_test.cu').read()
 
-- 
cgit