Random generators

MonteCarlo/CTG.cpp

@@ -0,0 +1,67 @@
+#include <fstream>
+
+#include "MonteCarloTests.h"
+#include "Common.h"
+
+CTG::CTG(size_t _randomNumbers, size_t _threadCount)
+{
+	randomNumbers = _randomNumbers;
+	threadCount = _threadCount;
+	randomBuffer = std::vector<float>(threadCount * randomNumbers);
+	seedBuffer = std::vector<float>(threadCount);
+	kernel_name = "randomCTG";
+
+	for (int i = 0; i < threadCount; ++i)
+	{
+		seedBuffer[i] = rand();
+	}
+}
+
+void CTG::collect_results(cl::CommandQueue* queue) {
+	
+}
+
+void CTG::gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event) {
+	cl_int err = CL_SUCCESS;
+	cl::Kernel kernel = cl::Kernel(*program, kernel_name.c_str(), &err);
+	CheckCLError(err);
+
+	clInputBuffer = cl::Buffer(*context, CL_MEM_READ_ONLY, sizeof(float) * threadCount, NULL, &err);
+	queue->enqueueWriteBuffer(clInputBuffer, true, 0, sizeof(float) * threadCount, seedBuffer.data());
+
+	clResultBuffer = cl::Buffer(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomNumbers, NULL, &err);
+
+	// Set the kernel parameters	
+	kernel.setArg(0, randomNumbers);
+	kernel.setArg(1, clInputBuffer);
+	kernel.setArg(2, clResultBuffer);
+
+	// Enqueue the kernel: threadCount threads in total, each generating random numbers in [0,1] randomNumbers times
+	queue->enqueueNDRangeKernel(kernel,
+		cl::NullRange,
+		cl::NDRange(threadCount, 1),
+		cl::NullRange,
+		NULL,
+		Event);
+}
+
+void CTG::cpu_compute()
+{
+}
+
+bool CTG::validate_results()
+{
+	std::ofstream ofs("randoms_" + description() + ".txt", std::ofstream::out);
+	for (int i = 0; i < threadCount; ++i) {
+		for (int j = 0; j < randomNumbers; ++j) {
+			ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
+		}
+	}
+	ofs.close();
+	return true;
+}
+
+std::string CTG::description()
+{
+	return kernel_name + "(numbers=" + std::to_string(randomNumbers) + ",threads=" + std::to_string(threadCount) + ")";
+}

MonteCarlo/Halton.cpp

@@ -0,0 +1,60 @@
+#include <fstream>
+
+#include "MonteCarloTests.h"
+#include "Common.h"
+
+Halton::Halton(size_t _randomNumbers, size_t _threadCount, size_t _base) {
+	randomNumbers = _randomNumbers;
+	threadCount = _threadCount;
+	base = _base;
+	randomBuffer = std::vector<float>(threadCount * randomNumbers);
+	kernel_name = "haltonSequence";
+}
+
+void Halton::collect_results(cl::CommandQueue* queue)
+{
+	queue->enqueueReadBuffer(clResultBuffer, true, 0, sizeof(float) * threadCount * randomNumbers, randomBuffer.data());
+}
+
+void Halton::gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event)
+{
+	cl_int err = CL_SUCCESS;
+	cl::Kernel kernel = cl::Kernel(*program, kernel_name.c_str(), &err);
+	CheckCLError(err);
+
+	clResultBuffer = cl::Buffer(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomNumbers, NULL, &err);
+
+	// Set the kernel parameters	
+	kernel.setArg(0, randomNumbers);
+	kernel.setArg(1, base);
+	kernel.setArg(2, clResultBuffer);
+
+	// Enqueue the kernel: threadCount threads in total, each generating random numbers in [0,1] randomNumbers times
+	queue->enqueueNDRangeKernel(kernel,
+		cl::NullRange,
+		cl::NDRange(threadCount, 1),
+		cl::NullRange,
+		NULL,
+		Event);
+}
+
+void Halton::cpu_compute()
+{
+}
+
+bool Halton::validate_results()
+{
+	std::ofstream ofs("randoms_" + description() + ".txt", std::ofstream::out);
+	for (int i = 0; i < threadCount; ++i) {
+		for (int j = 0; j < randomNumbers; ++j) {
+			ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
+		}
+	}
+	ofs.close();
+	return true;
+}
+
+std::string Halton::description()
+{
+	return kernel_name + "(numbers=" + std::to_string(randomNumbers) + ",threads=" + std::to_string(threadCount) + ",base=" + std::to_string(base) + ")";
+}

MonteCarlo/Hybrid.cpp

@@ -0,0 +1,69 @@
+#include <fstream>
+
+#include "MonteCarloTests.h"
+#include "Common.h"
+
+Hybrid::Hybrid(size_t _randomNumbers, size_t _threadCount)
+{
+	randomNumbers = _randomNumbers;
+	threadCount = _threadCount;
+	randomBuffer = std::vector<float>(threadCount * randomNumbers);
+	seedBuffer = std::vector<float>(threadCount);
+	kernel_name = "hybridRNG";
+
+	for (int i = 0; i < threadCount; ++i)
+	{
+		seedBuffer[i] = rand();
+	}
+}
+
+void Hybrid::collect_results(cl::CommandQueue* queue)
+{
+	queue->enqueueReadBuffer(clResultBuffer, true, 0, sizeof(float) * threadCount * randomNumbers, randomBuffer.data());
+}
+
+void Hybrid::gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event)
+{
+	cl_int err = CL_SUCCESS;
+	cl::Kernel kernel = cl::Kernel(*program, kernel_name.c_str(), &err);
+	CheckCLError(err);
+
+	clInputBuffer = cl::Buffer(*context, CL_MEM_READ_ONLY, sizeof(float) * threadCount, NULL, &err);
+	queue->enqueueWriteBuffer(clInputBuffer, true, 0, sizeof(float) * threadCount, seedBuffer.data());
+
+	clResultBuffer = cl::Buffer(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomNumbers, NULL, &err);
+
+	// Set the kernel parameters	
+	kernel.setArg(0, randomNumbers);
+	kernel.setArg(1, clInputBuffer);
+	kernel.setArg(2, clResultBuffer);
+
+	// Enqueue the kernel: threadCount threads in total, each generating random numbers in [0,1] randomNumbers times
+	queue->enqueueNDRangeKernel(kernel,
+		cl::NullRange,
+		cl::NDRange(threadCount, 1),
+		cl::NullRange,
+		NULL,
+		Event);
+}
+
+void Hybrid::cpu_compute()
+{
+}
+
+bool Hybrid::validate_results()
+{
+	std::ofstream ofs("randoms_" + description() + ".txt", std::ofstream::out);
+	for (int i = 0; i < threadCount; ++i) {
+		for (int j = 0; j < randomNumbers; ++j) {
+			ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
+		}
+	}
+	ofs.close();
+	return true;
+}
+
+std::string Hybrid::description()
+{
+	return kernel_name + "(numbers=" + std::to_string(randomNumbers) + ",threads=" + std::to_string(threadCount) + ")";
+}

MonteCarlo/LCG.cpp

@@ -0,0 +1,69 @@
+#include <fstream>
+
+#include "MonteCarloTests.h"
+#include "Common.h"
+
+LCG::LCG(size_t _randomNumbers, size_t _threadCount)
+{
+	randomNumbers = _randomNumbers;
+	threadCount = _threadCount;
+	randomBuffer = std::vector<float>(threadCount * randomNumbers);
+	seedBuffer = std::vector<float>(threadCount);
+	kernel_name = "randomLCG";
+
+	for (int i = 0; i < threadCount; ++i)
+	{
+		seedBuffer[i] = rand();
+	}
+}
+
+void LCG::collect_results(cl::CommandQueue* queue)
+{
+	queue->enqueueReadBuffer(clResultBuffer, true, 0, sizeof(float) * threadCount * randomNumbers, randomBuffer.data());
+}
+
+void LCG::gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event)
+{
+	cl_int err = CL_SUCCESS;
+	cl::Kernel kernel = cl::Kernel(*program, kernel_name.c_str(), &err);
+	CheckCLError(err);
+
+	clInputBuffer = cl::Buffer(*context, CL_MEM_READ_ONLY, sizeof(float) * threadCount, NULL, &err);
+	queue->enqueueWriteBuffer(clInputBuffer, true, 0, sizeof(float) * threadCount, seedBuffer.data());
+
+	clResultBuffer = cl::Buffer(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomNumbers, NULL, &err);
+
+	// Set the kernel parameters	
+	kernel.setArg(0, randomNumbers);
+	kernel.setArg(1, clInputBuffer);
+	kernel.setArg(2, clResultBuffer);
+
+	// Enqueue the kernel: threadCount threads in total, each generating random numbers in [0,1] randomNumbers times
+	queue->enqueueNDRangeKernel(kernel,
+		cl::NullRange,
+		cl::NDRange(threadCount, 1),
+		cl::NullRange,
+		NULL,
+		Event);
+}
+
+void LCG::cpu_compute()
+{
+}
+
+bool LCG::validate_results()
+{
+	std::ofstream ofs("randoms_" + description() + ".txt", std::ofstream::out);
+	for (int i = 0; i < threadCount; ++i) {
+		for (int j = 0; j < randomNumbers; ++j) {
+			ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
+		}
+	}
+	ofs.close();
+	return true;
+}
+
+std::string LCG::description()
+{
+	return kernel_name + "(numbers=" + std::to_string(randomNumbers) +",threads="+std::to_string(threadCount) + ")";
+}

MonteCarlo/LFG.cpp

@@ -0,0 +1,74 @@
+#include <fstream>
+
+#include "MonteCarloTests.h"
+#include "Common.h"
+
+LFG::LFG(size_t _randomNumbers, size_t _threadCount, size_t _randomStateSize)
+{
+	randomNumbers = _randomNumbers;
+	threadCount = _threadCount;
+	randomBuffer = std::vector<float>(threadCount * randomNumbers);
+	seedBuffer = std::vector<float>(threadCount);
+	kernel_name = "randomLFG";
+	randomStateSize = _randomStateSize;
+
+	for (int i = 0; i < threadCount; ++i)
+	{
+		seedBuffer[i] = rand();
+	}
+}
+
+void LFG::collect_results(cl::CommandQueue* queue)
+{
+	queue->enqueueReadBuffer(clResultBuffer, true, 0, sizeof(float) * threadCount * randomNumbers, randomBuffer.data());
+}
+
+void LFG::gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event)
+{
+	cl_int err = CL_SUCCESS;
+	cl::Kernel kernel = cl::Kernel(*program, kernel_name.c_str(), &err);
+	CheckCLError(err);
+
+	clInputBuffer = cl::Buffer(*context, CL_MEM_READ_ONLY, sizeof(float) * threadCount, NULL, &err);
+	queue->enqueueWriteBuffer(clInputBuffer, true, 0, sizeof(float) * threadCount, seedBuffer.data());
+
+	clResultBuffer = cl::Buffer(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomNumbers, NULL, &err);
+	cl::Buffer randomStates(*context, CL_MEM_WRITE_ONLY, sizeof(float) * threadCount * randomStateSize, NULL, &err);
+
+	// Set the kernel parameters	
+	kernel.setArg(0, randomNumbers);
+	kernel.setArg(1, clInputBuffer);
+	kernel.setArg(2, randomStateSize);
+	kernel.setArg(3, randomStates);
+	kernel.setArg(4, clResultBuffer);
+
+	// Enqueue the kernel: threadCount threads in total, each generating random numbers in [0,1] randomNumbers times
+	queue->enqueueNDRangeKernel(kernel,
+		cl::NullRange,
+		cl::NDRange(threadCount, 1),
+		cl::NullRange,
+		NULL,
+		Event);
+}
+
+void LFG::cpu_compute()
+{
+}
+
+bool LFG::validate_results()
+{
+	std::ofstream ofs("randoms_" + description() + ".txt", std::ofstream::out);
+	for (int i = 0; i < threadCount; ++i) {
+		for (int j = 0; j < randomNumbers; ++j) {
+			ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
+		}
+	}
+	ofs.close();
+	return true;
+}
+
+std::string LFG::description()
+{
+	return kernel_name + "(numbers=" + std::to_string(randomNumbers) + ",threads=" + std::to_string(threadCount) + ",state="+std::to_string(randomStateSize) + ")";
+}
+

MonteCarlo/MonteCarlo.cpp

@@ -11,95 +11,13 @@
 #include "cl.hpp"
 #include <iostream>
 #include <fstream>
+#include <OpenCLHandler.h>
+#include "MonteCarloTests.h"
 
 const bool writeOutRandoms = true;
 const size_t randomNumbers = 1024;
 const size_t threadCount = 512;
 
-void capi()
-{
-	// Get a platform ID
-	cl_platform_id platformID;
-	clGetPlatformIDs(1, &platformID, NULL);
-	
-	// Get a device ID
-	cl_device_id deviceID;
-	clGetDeviceIDs(platformID, CL_DEVICE_TYPE_GPU, 1, &deviceID, NULL);
-	
-	// Create a context
-	cl_context context;
-	cl_context_properties contextProperties[] =
-	{ CL_CONTEXT_PLATFORM, (cl_context_properties)platformID, 0 };
-	context = clCreateContext(contextProperties, 1, &deviceID, NULL, NULL, NULL);
-	
-	// Create a command queue
-	cl_command_queue queue;
-	queue = clCreateCommandQueue(context, deviceID, CL_QUEUE_PROFILING_ENABLE, NULL);
-	
-	// Create an OpenCL program
-	std::string source = FileToString("../kernels/montecarlo.cl");
-	const char* csource = source.c_str();
-	cl_program program = clCreateProgramWithSource(context, 1, &csource, NULL, NULL);
-	cl_int err = clBuildProgram(program, 1, &deviceID, NULL, NULL, NULL);
-	if (err != CL_SUCCESS)
-	{
-		cl_uint logLength;
-		clGetProgramBuildInfo(program, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &logLength);
-		char* log = new char[logLength];
-		clGetProgramBuildInfo(program, deviceID, CL_PROGRAM_BUILD_LOG, logLength, log, 0);
-		std::cout << log << std::endl;
-		delete[] log;
-		exit(-1);
-	}
-
-	// Get the kernel handle
-	cl_kernel kernel = clCreateKernel(program, "randomLCG", &err);
-	if(!CheckCLError(err)) exit(-1);
-
-	// Allocate memory for random numbers and random seeds
-	// Every thread receives a private seed, stored in seedBuffer/clSeedBuffer
-	// Every thread is supposed to generate randomNumbers random numbers
-	std::vector<float> randomBuffer(threadCount * randomNumbers);
-	std::vector<float> seedBuffer(threadCount);
-	for (int i = 0; i < threadCount; ++i)
-	{
-		seedBuffer[i] = rand();
-	}
-
-	cl_mem randomBufferDev;
-	randomBufferDev = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float)* threadCount * randomNumbers, NULL, &err);
-	if (!CheckCLError(err)) exit(-1);
-	cl_mem seedBufferDev;
-	seedBufferDev = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float) * threadCount, NULL, &err);
-	if (!CheckCLError(err)) exit(-1);
-
-	clEnqueueWriteBuffer(queue, seedBufferDev, CL_TRUE, 0, sizeof(float) * threadCount, seedBuffer.data(), 0, NULL, NULL);
-
-	// Set the kernel paramateres
-	clSetKernelArg(kernel, 0, sizeof(int), &randomNumbers);
-	clSetKernelArg(kernel, 1, sizeof(cl_mem), &seedBufferDev);
-	clSetKernelArg(kernel, 2, sizeof(cl_mem), &randomBufferDev);
-	// Enqueue the kernel: threadCount threads in total, each generating randomNumbers random numbers in [0,1]
-	clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &threadCount, NULL, 0, NULL, NULL);
-	// Copy the result back to the host
-	clEnqueueReadBuffer(queue, randomBufferDev, CL_TRUE, 0, sizeof(float) * threadCount * randomNumbers, randomBuffer.data(), 0, NULL, NULL);
-
-	// Write out the output
-	if(writeOutRandoms == true)
-	{
-		std::ofstream ofs("randoms.txt", std::ofstream::out);
-		for (int i = 0; i < threadCount; ++i) {
-			for (int j = 0; j < randomNumbers; ++j) {
-				ofs << j << " " << randomBuffer[i + j * threadCount] << std::endl;
-			}
-			ofs << std::endl;
-		}
-		ofs.close();
-	}
-
-	std::cout << "Finished" << std::endl;
-}
-
 void cppapi()
 {
 	cl_int err = CL_SUCCESS;
@@ -126,7 +44,7 @@ void cppapi()
 	cl::CommandQueue queue(context, devices[0], 0, &err);
 
 	// Create the OpenCL program
-	std::string programSource = FileToString("../kernels/programs.cl");
+	std::string programSource = FileToString("../kernels/montecarlo.cl");
 	cl::Program program = cl::Program(context, programSource);
 	err = program.build(devices);
 	if (!CheckCLError(err))
@@ -195,8 +113,14 @@ void cppapi()
 
 int main()
 {
-	capi();
-	cppapi();
+	//cppapi();
+	OpenCLHandler handler("../kernels/montecarlo.cl");
+
+	//handler.run_test(new LCG(randomNumbers, threadCount));
+	//handler.run_test(new LFG(randomNumbers, threadCount, 256));
+	//handler.run_test(new CTG(randomNumbers, threadCount));
+	//handler.run_test(new Hybrid(randomNumbers, threadCount));
+	handler.run_test(new Halton(randomNumbers, threadCount, 2));
 	return 0;
 }
 

MonteCarlo/MonteCarlo.vcxproj

@@ -72,6 +72,11 @@
     </Link>
   </ItemDefinitionGroup>
   <ItemGroup>
+    <ClCompile Include="CTG.cpp" />
+    <ClCompile Include="Halton.cpp" />
+    <ClCompile Include="Hybrid.cpp" />
+    <ClCompile Include="LCG.cpp" />
+    <ClCompile Include="LFG.cpp" />
     <ClCompile Include="MonteCarlo.cpp" />
   </ItemGroup>
   <ItemGroup>
@@ -83,6 +88,9 @@
       <Project>{f66311cb-c60d-43de-890c-7e6d8179ca44}</Project>
     </ProjectReference>
   </ItemGroup>
+  <ItemGroup>
+    <ClInclude Include="MonteCarloTests.h" />
+  </ItemGroup>
   <Import Project="$(VCTargetsPath)\Microsoft.Cpp.targets" />
   <ImportGroup Label="ExtensionTargets">
   </ImportGroup>

MonteCarlo/MonteCarlo.vcxproj.filters

@@ -21,6 +21,21 @@
     <ClCompile Include="MonteCarlo.cpp">
       <Filter>Source Files</Filter>
     </ClCompile>
+    <ClCompile Include="LCG.cpp">
+      <Filter>Source Files</Filter>
+    </ClCompile>
+    <ClCompile Include="LFG.cpp">
+      <Filter>Source Files</Filter>
+    </ClCompile>
+    <ClCompile Include="CTG.cpp">
+      <Filter>Source Files</Filter>
+    </ClCompile>
+    <ClCompile Include="Hybrid.cpp">
+      <Filter>Source Files</Filter>
+    </ClCompile>
+    <ClCompile Include="Halton.cpp">
+      <Filter>Source Files</Filter>
+    </ClCompile>
   </ItemGroup>
   <ItemGroup>
     <None Include="..\kernels\montecarlo.cl">
@@ -30,4 +45,9 @@
       <Filter>Kernels</Filter>
     </None>
   </ItemGroup>
+  <ItemGroup>
+    <ClInclude Include="MonteCarloTests.h">
+      <Filter>Header Files</Filter>
+    </ClInclude>
+  </ItemGroup>
 </Project>

MonteCarlo/MonteCarloTests.h

@@ -0,0 +1,84 @@
+#pragma once
+#include "Tests.h"
+
+
+class LCG : public TestCase {
+private:
+	size_t randomNumbers;
+	size_t threadCount;
+	std::vector<float> randomBuffer;
+	std::vector<float> seedBuffer;
+	std::string kernel_name;
+public:
+	LCG(size_t _randomNumbers, size_t _threadCount);
+	void collect_results(cl::CommandQueue* queue);
+	void gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event);
+	void cpu_compute();
+	bool validate_results();
+	std::string description();
+};
+
+class LFG : public TestCase {
+private:
+	size_t randomNumbers;
+	size_t threadCount;
+	std::vector<float> randomBuffer;
+	std::vector<float> seedBuffer;
+	std::string kernel_name;
+	size_t randomStateSize;
+public:
+	LFG(size_t _randomNumbers, size_t _threadCount, size_t _randomStateSize);
+	void collect_results(cl::CommandQueue* queue);
+	void gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event);
+	void cpu_compute();
+	bool validate_results();
+	std::string description();
+};
+
+class CTG : public TestCase {
+private:
+	size_t randomNumbers;
+	size_t threadCount;
+	std::vector<float> randomBuffer;
+	std::vector<float> seedBuffer;
+	std::string kernel_name;
+public:
+	CTG(size_t _randomNumbers, size_t _threadCount);
+	void collect_results(cl::CommandQueue* queue);
+	void gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event);
+	void cpu_compute();
+	bool validate_results();
+	std::string description();
+};
+
+class Hybrid : public TestCase {
+private:
+	size_t randomNumbers;
+	size_t threadCount;
+	std::vector<float> randomBuffer;
+	std::vector<float> seedBuffer;
+	std::string kernel_name;
+public:
+	Hybrid(size_t _randomNumbers, size_t _threadCount);
+	void collect_results(cl::CommandQueue* queue);
+	void gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event);
+	void cpu_compute();
+	bool validate_results();
+	std::string description();
+};
+
+class Halton : public TestCase {
+private:
+	size_t randomNumbers;
+	size_t threadCount;
+	size_t base;
+	std::vector<float> randomBuffer;
+	std::string kernel_name;
+public:
+	Halton(size_t _randomNumbers, size_t _threadCount, size_t _base);
+	void collect_results(cl::CommandQueue* queue);
+	void gpu_compute(cl::Context* context, cl::CommandQueue* queue, cl::Program* program, cl::Event* Event);
+	void cpu_compute();
+	bool validate_results();
+	std::string description();
+};