test for mu cuda class

posts/cmake/CMakeLists.txt

@@ -0,0 +1,46 @@
+
+cmake_minimum_required(VERSION 3.8 FATAL_ERROR)
+project(cmake_and_cuda LANGUAGES CXX CUDA)
+
+include(CTest)
+
+add_library(particles STATIC
+  randomize.cpp
+  randomize.h
+  particle.cu
+  particle.h
+  v3.cu
+  v3.h
+  )
+
+# Request that particles be built with -std=c++11
+# As this is a public compile feature anything that links to particles
+# will also build with -std=c++11
+target_compile_features(particles PUBLIC cxx_std_11)
+
+# We need to explicitly state that we need all CUDA files in the particle
+# library to be built with -dc as the member functions could be called by
+# other libraries and executables
+set_target_properties( particles
+                       PROPERTIES CUDA_SEPARABLE_COMPILATION ON
+                      )
+
+if(BUILD_TESTING)
+
+  add_executable(particle_test test.cu)
+
+  set_target_properties(particle_test PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
+  target_link_libraries(particle_test PRIVATE particles)
+
+  add_test(NAME particles_10k COMMAND particle_test 10000 )
+  add_test(NAME particles_256k COMMAND particle_test 256000 )
+
+  if(APPLE)
+    # We need to add the default path to the driver (libcuda.dylib) as an rpath,
+    # so that the static cuda runtime can find it at runtime.
+    set_property(TARGET particle_test PROPERTY BUILD_RPATH ${CMAKE_CUDA_IMPLICIT_LINK_DIRECTORIES})
+  endif()
+endif()
+
+
+

posts/cmake/particle.cu

@@ -0,0 +1,50 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+#include "particle.h"
+
+particle::particle() : 	position(), velocity(), totalDistance(1,0,0)
+{
+}
+
+__device__ __host__
+void particle::advance(float d)
+{
+  velocity.normalize();
+  auto dx = d * velocity.x;
+  position.x += dx;
+  totalDistance.x += dx;
+  auto dy = d * velocity.y;
+  position.y += dy;
+  totalDistance.y += dy;
+  auto dz = d * velocity.z;
+  position.z += dz;
+  totalDistance.z += dz;
+  velocity.scramble();
+}
+
+const v3& particle::getTotalDistance() const
+{	return totalDistance; }

posts/cmake/particle.h

@@ -0,0 +1,46 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+#ifndef __particle_h__
+#define __particle_h__
+
+#include "v3.h"
+
+class particle
+{
+	private:
+		v3 position;
+		v3 velocity;
+		v3 totalDistance;
+
+	public:
+		particle();
+		__host__ __device__ void advance(float dist);
+		const v3& getTotalDistance() const;
+
+};
+
+#endif

posts/cmake/randomize.cpp

@@ -0,0 +1,42 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+
+#include <random>
+
+namespace {
+  std::random_device randDevice;
+  std::minstd_rand engine(randDevice());
+  std::normal_distribution<float> normalDist(0, RAND_MAX);
+}
+
+
+void randomize(float& x, float& y, float& z)
+{
+  x = normalDist(engine);
+  y = normalDist(engine);
+  z = normalDist(engine);
+}

posts/cmake/randomize.h

@@ -0,0 +1,32 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+#ifndef __randomize_h__
+#define __randomize_h__
+
+void randomize(float& x, float& y, float& z);
+
+#endif

posts/cmake/test.cu

@@ -0,0 +1,103 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+#include "particle.h"
+#include <stdlib.h>
+#include <stdio.h>
+
+__global__ void advanceParticles(float dt, particle * pArray, int nParticles)
+{
+	int idx = threadIdx.x + blockIdx.x*blockDim.x;
+	if(idx < nParticles)
+	{
+		pArray[idx].advance(dt);
+	}
+}
+
+int main(int argc, char ** argv)
+{
+	cudaError_t error;
+	int n = 1000000;
+	if(argc > 1)	{ n = atoi(argv[1]);}     // Number of particles
+	if(argc > 2)	{	srand(atoi(argv[2])); } // Random seed
+
+	error = cudaGetLastError();
+	if (error != cudaSuccess)
+  	{
+  	printf("0 %s\n",cudaGetErrorString(error));
+  	exit(1);
+  	}
+
+	particle * pArray = new particle[n];
+	particle * devPArray = NULL;
+	cudaMalloc(&devPArray, n*sizeof(particle));
+	cudaDeviceSynchronize(); error = cudaGetLastError();
+	if (error != cudaSuccess)
+  	{
+  	printf("1 %s\n",cudaGetErrorString(error));
+  	exit(1);
+  	}
+
+	cudaMemcpy(devPArray, pArray, n*sizeof(particle), cudaMemcpyHostToDevice);
+	cudaDeviceSynchronize(); error = cudaGetLastError();
+	if (error != cudaSuccess)
+  	{
+  	printf("2 %s\n",cudaGetErrorString(error));
+  	exit(1);
+  	}
+
+	for(int i=0; i<100; i++)
+	{
+		float dt = (float)rand()/(float) RAND_MAX; // Random distance each step
+		advanceParticles<<< 1 +  n/256, 256>>>(dt, devPArray, n);
+		error = cudaGetLastError();
+		if (error != cudaSuccess)
+    	{
+    	printf("3 %s\n",cudaGetErrorString(error));
+    	exit(1);
+    	}
+
+		cudaDeviceSynchronize();
+	}
+	cudaMemcpy(pArray, devPArray, n*sizeof(particle), cudaMemcpyDeviceToHost);
+
+	v3 totalDistance(0,0,0);
+	v3 temp;
+	for(int i=0; i<n; i++)
+	{
+		temp = pArray[i].getTotalDistance();
+		totalDistance.x += temp.x;
+		totalDistance.y += temp.y;
+		totalDistance.z += temp.z;
+	}
+	float avgX = totalDistance.x /(float)n;
+	float avgY = totalDistance.y /(float)n;
+	float avgZ = totalDistance.z /(float)n;
+	float avgNorm = sqrt(avgX*avgX + avgY*avgY + avgZ*avgZ);
+	printf(	"Moved %d particles 100 steps. Average distance traveled is |(%f, %f, %f)| = %f\n",
+					n, avgX, avgY, avgZ, avgNorm);
+	return 0;
+}

posts/cmake/v3.cu

@@ -0,0 +1,60 @@
+/* Copyright (c) 1993-2015, NVIDIA CORPORATION. All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions
+ * are met:
+ *  * Redistributions of source code must retain the above copyright
+ *    notice, this list of conditions and the following disclaimer.
+ *  * Redistributions in binary form must reproduce the above copyright
+ *    notice, this list of conditions and the following disclaimer in the
+ *    documentation and/or other materials provided with the distribution.
+ *  * Neither the name of NVIDIA CORPORATION nor the names of its
+ *    contributors may be used to endorse or promote products derived
+ *    from this software without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
+ * EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
+ * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED.  IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
+ * OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
+ * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
+ * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+#include "v3.h"
+#include "randomize.h"
+#include <math.h>
+
+v3::v3()
+{
+	::randomize(x, y, z);
+}
+
+v3::v3(float xIn, float yIn, float zIn) : x(xIn), y(yIn), z(zIn)
+{}
+
+void v3::randomize()
+{
+	::randomize(x, y, z);
+}
+
+__host__ __device__ void v3::normalize()
+{
+	float t = sqrt(x*x + y*y + z*z);
+	x /= t;
+	y /= t;
+	z /= t;
+}
+
+__host__ __device__ void v3::scramble()
+{
+	float tx = 0.317f*(x + 1.0) + y + z * x * x + y + z;
+	float ty = 0.619f*(y + 1.0) + y * y + x * y * z + y + x;
+	float tz = 0.124f*(z + 1.0) + z * y + x * y * z + y + x;
+	x = tx;
+	y = ty;
+	z = tz;
+}