Rewrote the entire thing using my new OpenCL template. Makes small one offs like these reallly simple
parent
08fd8b2398
commit
7743321465
@ -1,63 +0,0 @@
|
||||
###############################################################################
|
||||
# Set default behavior to automatically normalize line endings.
|
||||
###############################################################################
|
||||
* text=auto
|
||||
|
||||
###############################################################################
|
||||
# Set default behavior for command prompt diff.
|
||||
#
|
||||
# This is need for earlier builds of msysgit that does not have it on by
|
||||
# default for csharp files.
|
||||
# Note: This is only used by command line
|
||||
###############################################################################
|
||||
#*.cs diff=csharp
|
||||
|
||||
###############################################################################
|
||||
# Set the merge driver for project and solution files
|
||||
#
|
||||
# Merging from the command prompt will add diff markers to the files if there
|
||||
# are conflicts (Merging from VS is not affected by the settings below, in VS
|
||||
# the diff markers are never inserted). Diff markers may cause the following
|
||||
# file extensions to fail to load in VS. An alternative would be to treat
|
||||
# these files as binary and thus will always conflict and require user
|
||||
# intervention with every merge. To do so, just uncomment the entries below
|
||||
###############################################################################
|
||||
#*.sln merge=binary
|
||||
#*.csproj merge=binary
|
||||
#*.vbproj merge=binary
|
||||
#*.vcxproj merge=binary
|
||||
#*.vcproj merge=binary
|
||||
#*.dbproj merge=binary
|
||||
#*.fsproj merge=binary
|
||||
#*.lsproj merge=binary
|
||||
#*.wixproj merge=binary
|
||||
#*.modelproj merge=binary
|
||||
#*.sqlproj merge=binary
|
||||
#*.wwaproj merge=binary
|
||||
|
||||
###############################################################################
|
||||
# behavior for image files
|
||||
#
|
||||
# image files are treated as binary by default.
|
||||
###############################################################################
|
||||
#*.jpg binary
|
||||
#*.png binary
|
||||
#*.gif binary
|
||||
|
||||
###############################################################################
|
||||
# diff behavior for common document formats
|
||||
#
|
||||
# Convert binary document formats to text before diffing them. This feature
|
||||
# is only available from the command line. Turn it on by uncommenting the
|
||||
# entries below.
|
||||
###############################################################################
|
||||
#*.doc diff=astextplain
|
||||
#*.DOC diff=astextplain
|
||||
#*.docx diff=astextplain
|
||||
#*.DOCX diff=astextplain
|
||||
#*.dot diff=astextplain
|
||||
#*.DOT diff=astextplain
|
||||
#*.pdf diff=astextplain
|
||||
#*.PDF diff=astextplain
|
||||
#*.rtf diff=astextplain
|
||||
#*.RTF diff=astextplain
|
@ -1,212 +1,47 @@
|
||||
## Ignore Visual Studio temporary files, build results, and
|
||||
## files generated by popular Visual Studio add-ons.
|
||||
|
||||
# User-specific files
|
||||
*.suo
|
||||
*.user
|
||||
*.userosscache
|
||||
*.sln.docstates
|
||||
|
||||
# User-specific files (MonoDevelop/Xamarin Studio)
|
||||
*.userprefs
|
||||
|
||||
# Build results
|
||||
[Dd]ebug/
|
||||
[Dd]ebugPublic/
|
||||
[Rr]elease/
|
||||
[Rr]eleases/
|
||||
x64/
|
||||
x86/
|
||||
build/
|
||||
bld/
|
||||
[Bb]in/
|
||||
[Oo]bj/
|
||||
|
||||
# Visual Studio 2015 cache/options directory
|
||||
.vs/
|
||||
|
||||
# MSTest test Results
|
||||
[Tt]est[Rr]esult*/
|
||||
[Bb]uild[Ll]og.*
|
||||
|
||||
# NUNIT
|
||||
*.VisualState.xml
|
||||
TestResult.xml
|
||||
|
||||
# Build Results of an ATL Project
|
||||
[Dd]ebugPS/
|
||||
[Rr]eleasePS/
|
||||
dlldata.c
|
||||
|
||||
# DNX
|
||||
project.lock.json
|
||||
artifacts/
|
||||
|
||||
*_i.c
|
||||
*_p.c
|
||||
*_i.h
|
||||
*.ilk
|
||||
*.meta
|
||||
*.obj
|
||||
*.pch
|
||||
*.pdb
|
||||
*.pgc
|
||||
*.pgd
|
||||
*.rsp
|
||||
*.sbr
|
||||
*.tlb
|
||||
*.tli
|
||||
*.tlh
|
||||
*.tmp
|
||||
*.tmp_proj
|
||||
# Compiled source #
|
||||
###################
|
||||
*.com
|
||||
*.class
|
||||
*.dll
|
||||
*.exe
|
||||
*.o
|
||||
*.so
|
||||
|
||||
# Packages #
|
||||
############
|
||||
# it's better to unpack these files and commit the raw source
|
||||
# git has its own built in compression methods
|
||||
*.7z
|
||||
*.dmg
|
||||
*.gz
|
||||
*.iso
|
||||
*.jar
|
||||
*.rar
|
||||
*.tar
|
||||
*.zip
|
||||
|
||||
# Logs and databases #
|
||||
######################
|
||||
*.log
|
||||
*.vspscc
|
||||
*.vssscc
|
||||
.builds
|
||||
*.pidb
|
||||
*.svclog
|
||||
*.scc
|
||||
|
||||
# Chutzpah Test files
|
||||
_Chutzpah*
|
||||
|
||||
# Visual C++ cache files
|
||||
ipch/
|
||||
*.aps
|
||||
*.ncb
|
||||
*.opensdf
|
||||
*.sdf
|
||||
*.cachefile
|
||||
|
||||
# Visual Studio profiler
|
||||
*.psess
|
||||
*.vsp
|
||||
*.vspx
|
||||
|
||||
# TFS 2012 Local Workspace
|
||||
$tf/
|
||||
|
||||
# Guidance Automation Toolkit
|
||||
*.gpState
|
||||
|
||||
# ReSharper is a .NET coding add-in
|
||||
_ReSharper*/
|
||||
*.[Rr]e[Ss]harper
|
||||
*.DotSettings.user
|
||||
|
||||
# JustCode is a .NET coding add-in
|
||||
.JustCode
|
||||
|
||||
# TeamCity is a build add-in
|
||||
_TeamCity*
|
||||
|
||||
# DotCover is a Code Coverage Tool
|
||||
*.dotCover
|
||||
|
||||
# NCrunch
|
||||
_NCrunch_*
|
||||
.*crunch*.local.xml
|
||||
|
||||
# MightyMoose
|
||||
*.mm.*
|
||||
AutoTest.Net/
|
||||
|
||||
# Web workbench (sass)
|
||||
.sass-cache/
|
||||
|
||||
# Installshield output folder
|
||||
[Ee]xpress/
|
||||
|
||||
# DocProject is a documentation generator add-in
|
||||
DocProject/buildhelp/
|
||||
DocProject/Help/*.HxT
|
||||
DocProject/Help/*.HxC
|
||||
DocProject/Help/*.hhc
|
||||
DocProject/Help/*.hhk
|
||||
DocProject/Help/*.hhp
|
||||
DocProject/Help/Html2
|
||||
DocProject/Help/html
|
||||
|
||||
# Click-Once directory
|
||||
publish/
|
||||
|
||||
# Publish Web Output
|
||||
*.[Pp]ublish.xml
|
||||
*.azurePubxml
|
||||
## TODO: Comment the next line if you want to checkin your
|
||||
## web deploy settings but do note that will include unencrypted
|
||||
## passwords
|
||||
#*.pubxml
|
||||
|
||||
*.publishproj
|
||||
|
||||
# NuGet Packages
|
||||
*.nupkg
|
||||
# The packages folder can be ignored because of Package Restore
|
||||
**/packages/*
|
||||
# except build/, which is used as an MSBuild target.
|
||||
!**/packages/build/
|
||||
# Uncomment if necessary however generally it will be regenerated when needed
|
||||
#!**/packages/repositories.config
|
||||
|
||||
# Windows Azure Build Output
|
||||
csx/
|
||||
*.build.csdef
|
||||
|
||||
# Windows Store app package directory
|
||||
AppPackages/
|
||||
|
||||
# Visual Studio cache files
|
||||
# files ending in .cache can be ignored
|
||||
*.[Cc]ache
|
||||
# but keep track of directories ending in .cache
|
||||
!*.[Cc]ache/
|
||||
|
||||
# Others
|
||||
ClientBin/
|
||||
[Ss]tyle[Cc]op.*
|
||||
~$*
|
||||
*~
|
||||
*.dbmdl
|
||||
*.dbproj.schemaview
|
||||
*.pfx
|
||||
*.publishsettings
|
||||
node_modules/
|
||||
orleans.codegen.cs
|
||||
|
||||
# RIA/Silverlight projects
|
||||
Generated_Code/
|
||||
|
||||
# Backup & report files from converting an old project file
|
||||
# to a newer Visual Studio version. Backup files are not needed,
|
||||
# because we have git ;-)
|
||||
_UpgradeReport_Files/
|
||||
Backup*/
|
||||
UpgradeLog*.XML
|
||||
UpgradeLog*.htm
|
||||
|
||||
# SQL Server files
|
||||
*.mdf
|
||||
*.ldf
|
||||
|
||||
# Business Intelligence projects
|
||||
*.rdl.data
|
||||
*.bim.layout
|
||||
*.bim_*.settings
|
||||
|
||||
# Microsoft Fakes
|
||||
FakesAssemblies/
|
||||
|
||||
# Node.js Tools for Visual Studio
|
||||
.ntvs_analysis.dat
|
||||
|
||||
# Visual Studio 6 build log
|
||||
*.plg
|
||||
|
||||
# Visual Studio 6 workspace options file
|
||||
*.opt
|
||||
|
||||
# LightSwitch generated files
|
||||
GeneratedArtifacts/
|
||||
_Pvt_Extensions/
|
||||
ModelManifest.xml
|
||||
*.sql
|
||||
*.sqlite
|
||||
|
||||
# OS generated files #
|
||||
######################
|
||||
.DS_Store
|
||||
.DS_Store?
|
||||
._*
|
||||
.Spotlight-V100
|
||||
.Trashes
|
||||
ehthumbs.db
|
||||
Thumbs.db
|
||||
|
||||
# CMake Generated Files #
|
||||
#########################
|
||||
CMakeCache.txt
|
||||
CMakeFiles
|
||||
CMakeScripts
|
||||
Makefile
|
||||
cmake_install.cmake
|
||||
install_manifest.txt
|
||||
CTestTestfile.cmake
|
||||
|
@ -0,0 +1,101 @@
|
||||
# Check versions
|
||||
message(STATUS "CMake version: ${CMAKE_VERSION}")
|
||||
cmake_minimum_required(VERSION 3.1)
|
||||
|
||||
set_property(GLOBAL PROPERTY USE_FOLDERS ON)
|
||||
|
||||
# Set the project name
|
||||
set(PNAME Conways)
|
||||
project(${PNAME})
|
||||
|
||||
# Set up variables, and find SFML
|
||||
if (WIN32)
|
||||
set(SFML_ROOT root CACHE STRING "User specified path")
|
||||
set(SFML_INCLUDE_DIR ${SFML_ROOT}/include)
|
||||
endif()
|
||||
|
||||
set(SFML_COMPONENTS graphics window system network audio)
|
||||
set(CMAKE_MODULE_PATH ${CMAKE_CURRENT_SOURCE_DIR})
|
||||
find_package(SFML 2.1 COMPONENTS ${SFML_COMPONENTS} REQUIRED)
|
||||
message(STATUS "SFML found: ${SFML_FOUND}")
|
||||
|
||||
# Find OpenCL
|
||||
find_package( OpenCL REQUIRED )
|
||||
message(STATUS "OpenCL found: ${OPENCL_FOUND}")
|
||||
|
||||
# Find OpenGL
|
||||
find_package( OpenGL REQUIRED)
|
||||
message(STATUS "OpenGL found: ${OPENGL_FOUND}")
|
||||
|
||||
# Include the directories for the main program, GL, CL and SFML's headers
|
||||
include_directories(${SFML_INCLUDE_DIR})
|
||||
include_directories(${OpenCL_INCLUDE_DIRS})
|
||||
include_directories(include)
|
||||
|
||||
# Glob all thr sources into their values
|
||||
file(GLOB_RECURSE SOURCES "src/*.cpp")
|
||||
file(GLOB_RECURSE HEADERS "include/*.h" "include/*.hpp")
|
||||
file(GLOB_RECURSE KERNELS "kernels/*.cl")
|
||||
|
||||
add_executable(${PNAME} ${SOURCES} ${HEADERS} ${KERNELS})
|
||||
|
||||
# Follow the sub directory structure to add sub-filters in VS
|
||||
# Gotta do it one by one unfortunately
|
||||
|
||||
foreach (source IN ITEMS ${SOURCES})
|
||||
if (IS_ABSOLUTE "${source}")
|
||||
|
||||
get_filename_component(filename ${source} DIRECTORY)
|
||||
|
||||
STRING(REGEX REPLACE "/" "\\\\" filename ${filename})
|
||||
|
||||
string(REGEX MATCHALL "src(.*)" substrings ${filename})
|
||||
list(GET substrings 0 substring)
|
||||
|
||||
SOURCE_GROUP(${substring} FILES ${source})
|
||||
|
||||
endif()
|
||||
endforeach()
|
||||
|
||||
foreach (source IN ITEMS ${HEADERS})
|
||||
if (IS_ABSOLUTE "${source}")
|
||||
|
||||
get_filename_component(filename ${source} DIRECTORY)
|
||||
|
||||
STRING(REGEX REPLACE "/" "\\\\" filename ${filename})
|
||||
|
||||
string(REGEX MATCHALL "include(.*)" substrings ${filename})
|
||||
list(GET substrings 0 substring)
|
||||
|
||||
SOURCE_GROUP(${substring} FILES ${source})
|
||||
|
||||
endif()
|
||||
endforeach()
|
||||
|
||||
foreach (source IN ITEMS ${KERNELS})
|
||||
if (IS_ABSOLUTE "${source}")
|
||||
|
||||
get_filename_component(filename ${source} DIRECTORY)
|
||||
|
||||
STRING(REGEX REPLACE "/" "\\\\" filename ${filename})
|
||||
|
||||
string(REGEX MATCHALL "kernels(.*)" substrings ${filename})
|
||||
list(GET substrings 0 substring)
|
||||
|
||||
SOURCE_GROUP(${substring} FILES ${source})
|
||||
|
||||
endif()
|
||||
endforeach()
|
||||
|
||||
# Link CL, GL, and SFML
|
||||
target_link_libraries (${PNAME} ${SFML_LIBRARIES} ${SFML_DEPENDENCIES})
|
||||
target_link_libraries (${PNAME} ${OpenCL_LIBRARY})
|
||||
target_link_libraries (${PNAME} ${OPENGL_LIBRARIES})
|
||||
|
||||
if (NOT WIN32)
|
||||
target_link_libraries (${PNAME} -lpthread)
|
||||
endif()
|
||||
|
||||
# Setup to use C++14
|
||||
set_property(TARGET ${PNAME} PROPERTY CXX_STANDARD 14)
|
||||
|
@ -1,395 +0,0 @@
|
||||
#include <CL/cl.h>
|
||||
#include <CL/opencl.h>
|
||||
#include <string.h>
|
||||
#include <stdio.h>
|
||||
#include <stdlib.h>
|
||||
#include <iostream>
|
||||
#include <string>
|
||||
#include <fstream>
|
||||
#include <random>
|
||||
#include <ctime>
|
||||
#include <SFML/Graphics.hpp>
|
||||
#include <windows.h>
|
||||
#include <GL/glew.h>
|
||||
#include <GLFW/glfw3.h>
|
||||
#include "Shader.hpp"
|
||||
|
||||
#define SUCCESS 0
|
||||
#define FAILURE 1
|
||||
|
||||
float elap_time() {
|
||||
static __int64 start = 0;
|
||||
static __int64 frequency = 0;
|
||||
|
||||
if (start == 0) {
|
||||
QueryPerformanceCounter((LARGE_INTEGER*)&start);
|
||||
QueryPerformanceFrequency((LARGE_INTEGER*)&frequency);
|
||||
return 0.0f;
|
||||
}
|
||||
|
||||
__int64 counter = 0;
|
||||
QueryPerformanceCounter((LARGE_INTEGER*)&counter);
|
||||
return (float)((counter - start) / double(frequency));
|
||||
}
|
||||
|
||||
// convert the kernel file into a string
|
||||
int convertToString(const char *filename, std::string& s)
|
||||
{
|
||||
size_t size;
|
||||
char* str;
|
||||
std::fstream f(filename, (std::fstream::in | std::fstream::binary));
|
||||
|
||||
if(f.is_open())
|
||||
{
|
||||
size_t fileSize;
|
||||
f.seekg(0, std::fstream::end);
|
||||
size = fileSize = (size_t)f.tellg();
|
||||
f.seekg(0, std::fstream::beg);
|
||||
str = new char[size+1];
|
||||
if(!str)
|
||||
{
|
||||
f.close();
|
||||
return 0;
|
||||
}
|
||||
|
||||
f.read(str, fileSize);
|
||||
f.close();
|
||||
str[size] = '\0';
|
||||
s = str;
|
||||
delete[] str;
|
||||
return 0;
|
||||
}
|
||||
std::cout << "Error: failed to open file\n:" << filename << std::endl;
|
||||
return FAILURE;
|
||||
}
|
||||
|
||||
void key_callback(GLFWwindow* window, int key, int scancode, int action, int mode) {
|
||||
if (key == GLFW_KEY_ESCAPE && action == GLFW_PRESS)
|
||||
glfwSetWindowShouldClose(window, GL_TRUE);
|
||||
}
|
||||
|
||||
int main(int argc, char* argv[])
|
||||
{
|
||||
int WINDOW_X = 1000;
|
||||
int WINDOW_Y = 1000;
|
||||
int GRID_WIDTH = WINDOW_X;
|
||||
int GRID_HEIGHT = WINDOW_Y;
|
||||
int WORKER_SIZE = 1000;
|
||||
|
||||
// ======================================= Setup OpenGL =======================================================
|
||||
|
||||
glfwInit();
|
||||
|
||||
glfwWindowHint(GLFW_CONTEXT_VERSION_MAJOR, 3);
|
||||
glfwWindowHint(GLFW_CONTEXT_VERSION_MINOR, 3);
|
||||
glfwWindowHint(GLFW_OPENGL_PROFILE, GLFW_OPENGL_CORE_PROFILE);
|
||||
glfwWindowHint(GLFW_RESIZABLE, GL_FALSE);
|
||||
|
||||
GLFWwindow* gl_window = glfwCreateWindow(GRID_WIDTH, GRID_HEIGHT, "GPU accelerated life", nullptr, nullptr);
|
||||
glfwMakeContextCurrent(gl_window);
|
||||
|
||||
glfwSetKeyCallback(gl_window, key_callback);
|
||||
|
||||
glewExperimental = GL_TRUE;
|
||||
glewInit();
|
||||
|
||||
glViewport(0, 0, GRID_WIDTH, GRID_HEIGHT);
|
||||
|
||||
glClearColor(0.2f, 0.3f, 0.3f, 1.0f);
|
||||
|
||||
// Stolen from LearnOpenGL, aint got no time for that
|
||||
Shader vert_frag_shaders("Z:\\VS_Projects\\Conway_OpenCL\\Conway_OpenCL\\vertex_shader.sh", "Z:\\VS_Projects\\Conway_OpenCL\\Conway_OpenCL\\fragment_shader.sh");
|
||||
|
||||
GLfloat vertices[] = {
|
||||
// Positions // Colors // Texture Coords
|
||||
1.0f, 1.0f, 0.0f, 1.0f, 0.0f, 0.0f, 1.0f, 1.0f, // Top Right
|
||||
1.0f, -1.0f, 0.0f, 0.0f, 1.0f, 0.0f, 1.0f, 0.0f, // Bottom Right
|
||||
-1.0f, -1.0f, 0.0f, 0.0f, 0.0f, 1.0f, 0.0f, 0.0f, // Bottom Left
|
||||
-1.0f, 1.0f, 0.0f, 1.0f, 1.0f, 0.0f, 0.0f, 1.0f // Top Left
|
||||
};
|
||||
GLuint indices[] = {
|
||||
0, 1, 3, // First Triangle
|
||||
1, 2, 3 // Second Triangle
|
||||
};
|
||||
|
||||
GLuint VBO, VAO, EBO;
|
||||
glGenVertexArrays(1, &VAO);
|
||||
glGenBuffers(1, &VBO);
|
||||
glGenBuffers(1, &EBO);
|
||||
|
||||
// Bind the Vertex Array Object first, then bind and set vertex buffer(s) and attribute pointer(s).
|
||||
glBindVertexArray(VAO);
|
||||
|
||||
glBindBuffer(GL_ARRAY_BUFFER, VBO);
|
||||
glBufferData(GL_ARRAY_BUFFER, sizeof(vertices), vertices, GL_STATIC_DRAW);
|
||||
|
||||
glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, EBO);
|
||||
glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(indices), indices, GL_STATIC_DRAW);
|
||||
|
||||
|
||||
// Position attribute
|
||||
glVertexAttribPointer(0, 3, GL_FLOAT, GL_FALSE, 8 * sizeof(GLfloat), (GLvoid*)0);
|
||||
glEnableVertexAttribArray(0);
|
||||
// Color attribute
|
||||
glVertexAttribPointer(1, 3, GL_FLOAT, GL_FALSE, 8 * sizeof(GLfloat), (GLvoid*)(3 * sizeof(GLfloat)));
|
||||
glEnableVertexAttribArray(1);
|
||||
// TexCoord attribute
|
||||
glVertexAttribPointer(2, 2, GL_FLOAT, GL_FALSE, 8 * sizeof(GLfloat), (GLvoid*)(6 * sizeof(GLfloat)));
|
||||
glEnableVertexAttribArray(2);
|
||||
|
||||
glBindVertexArray(0); // Unbind VAO
|
||||
|
||||
// ============================== OpenCL Setup ==================================================================
|
||||
|
||||
// Get the platforms
|
||||
cl_uint numPlatforms;
|
||||
cl_platform_id platform = NULL;
|
||||
cl_int status = clGetPlatformIDs(0, NULL, &numPlatforms); // Retrieve the number of platforms
|
||||
if (status != CL_SUCCESS) {
|
||||
std::cout << "Error: Getting platforms!" << std::endl;
|
||||
return FAILURE;
|
||||
}
|
||||
|
||||
// Choose the first available platform
|
||||
if(numPlatforms > 0) {
|
||||
cl_platform_id* platforms = new cl_platform_id[numPlatforms];
|
||||
status = clGetPlatformIDs(numPlatforms, platforms, NULL); // Now populate the array with the platforms
|
||||
platform = platforms[0];
|
||||
delete platforms;
|
||||
}
|
||||
|
||||
cl_uint numDevices = 0;
|
||||
cl_device_id *devices;
|
||||
status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &numDevices);
|
||||
if (numDevices == 0) { //no GPU available.
|
||||
std::cout << "No GPU device available." << std::endl;
|
||||
std::cout << "Choose CPU as default device." << std::endl;
|
||||
status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_CPU, 0, NULL, &numDevices);
|
||||
devices = new cl_device_id[numDevices];
|
||||
status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_CPU, numDevices, devices, NULL);
|
||||
}
|
||||
else {
|
||||
devices = new cl_device_id[numDevices];
|
||||
status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, numDevices, devices, NULL);
|
||||
}
|
||||
|
||||
HGLRC hGLRC = wglGetCurrentContext();
|
||||
HDC hDC = wglGetCurrentDC();
|
||||
cl_context_properties cps[] ={CL_CONTEXT_PLATFORM, (cl_context_properties)platform, CL_GL_CONTEXT_KHR, (cl_context_properties)hGLRC, CL_WGL_HDC_KHR, (cl_context_properties)hDC, 0 };
|
||||
|
||||
|
||||
cl_context context = clCreateContext(cps, 1, devices,NULL,NULL,NULL);
|
||||
cl_command_queue commandQueue = clCreateCommandQueue(context, devices[0], 0, NULL);
|
||||
|
||||
|
||||
// ============================== Kernel Compilation, Setup ====================================================
|
||||
|
||||
// Read the kernel from the file to a string
|
||||
const char *compute_kernel_filename = "Z:\\VS_Projects\\Conway_OpenCL\\Conway_OpenCL\\conway_compute.cl";
|
||||
const char *align_kernel_filename = "Z:\\VS_Projects\\Conway_OpenCL\\Conway_OpenCL\\conway_align.cl";
|
||||
|
||||
std::string compute_kernel_string;
|
||||
std::string align_kernel_string;
|
||||
|
||||
convertToString(compute_kernel_filename, compute_kernel_string);
|
||||
convertToString(align_kernel_filename, align_kernel_string);
|
||||
|
||||
// Create a program with the source
|
||||
const char *compute_source = compute_kernel_string.c_str();
|
||||
const char *align_source = align_kernel_string.c_str();
|
||||
|
||||
size_t compute_source_size[] = {strlen(compute_source)};
|
||||
size_t align_source_size[] = { strlen(align_source) };
|
||||
|
||||
cl_program compute_program = clCreateProgramWithSource(context, 1, &compute_source, compute_source_size, NULL);
|
||||
cl_program align_program = clCreateProgramWithSource(context, 1, &align_source, align_source_size, NULL);
|
||||
|
||||
// Build the compute program
|
||||
status = clBuildProgram(compute_program, 1, devices, NULL, NULL, NULL);
|
||||
|
||||
if (status == CL_BUILD_PROGRAM_FAILURE) {
|
||||
|
||||
size_t log_size;
|
||||
clGetProgramBuildInfo(compute_program, devices[0], CL_PROGRAM_BUILD_LOG, 0, NULL, &log_size);
|
||||
char *log = new char[log_size];
|
||||
|
||||
clGetProgramBuildInfo(compute_program, devices[0], CL_PROGRAM_BUILD_LOG, log_size, log, NULL);
|
||||
|
||||
std::cout << log << std::endl;
|
||||
}
|
||||
|
||||
// Build the align program
|
||||
status = clBuildProgram(align_program, 1, devices, NULL, NULL, NULL);
|
||||
|
||||
if (status == CL_BUILD_PROGRAM_FAILURE) {
|
||||
|
||||
size_t log_size;
|
||||
clGetProgramBuildInfo(align_program, devices[0], CL_PROGRAM_BUILD_LOG, 0, NULL, &log_size);
|
||||
char *log = new char[log_size];
|
||||
|
||||
clGetProgramBuildInfo(align_program, devices[0], CL_PROGRAM_BUILD_LOG, log_size, log, NULL);
|
||||
|
||||
std::cout << log << std::endl;
|
||||
}
|
||||
|
||||
// Now create the kernels
|
||||
cl_kernel compute_kernel = clCreateKernel(compute_program, "conway_compute", NULL);
|
||||
cl_kernel align_kernel = clCreateKernel(align_program, "conway_align", NULL);
|
||||
|
||||
|
||||
|
||||
// ======================================= Setup grid =========================================================
|
||||
|
||||
// Setup the rng
|
||||
std::mt19937 rng(time(NULL));
|
||||
std::uniform_int_distribution<int> rgen(0, 4); // 25% chance
|
||||
|
||||
// Init the grids
|
||||
unsigned char* node_grid = new unsigned char[GRID_WIDTH * GRID_HEIGHT];
|
||||
|
||||
for (int i = 0; i < GRID_WIDTH * GRID_HEIGHT; i++) {
|
||||
if (rgen(rng) == 1) {
|
||||
node_grid[i] = 1;
|
||||
}
|
||||
else {
|
||||
node_grid[i] = 0;
|
||||
}
|
||||
}
|
||||
|
||||
// ====================================== Setup Rendering ==========================================================
|
||||
|
||||
unsigned char* pixel_array = new sf::Uint8[WINDOW_X * WINDOW_Y * 4];
|
||||
|
||||
for (int i = 0; i < GRID_WIDTH * GRID_HEIGHT * 4; i += 4) {
|
||||
|
||||
pixel_array[i] = i % 255; // R?
|
||||
pixel_array[i + 1] = 70; // G?
|
||||
pixel_array[i + 2] = 100; // B?
|
||||
pixel_array[i + 3] = 100; // A?
|
||||
}
|
||||
|
||||
GLuint texture;
|
||||
|
||||
glGenTextures(1, &texture);
|
||||
glBindTexture(GL_TEXTURE_2D, texture);
|
||||
|
||||
//////////////////////////
|
||||
|
||||
glTexEnvf(GL_TEXTURE_ENV, GL_TEXTURE_ENV_MODE, GL_REPLACE);
|
||||
|
||||
glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_DECAL);
|
||||
glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_DECAL);
|
||||
|
||||
// when texture area is small, bilinear filter the closest mipmap
|
||||
glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR_MIPMAP_NEAREST);
|
||||
// when texture area is large, bilinear filter the first mipmap
|
||||
glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
|
||||
|
||||
// texture should tile
|
||||
glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
|
||||
glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
|
||||
|
||||
glTexImage2D(GL_TEXTURE_2D, 0, GL_RGBA, GRID_WIDTH, GRID_HEIGHT, 0, GL_RGBA, GL_UNSIGNED_BYTE, pixel_array);
|
||||
|
||||
glGenerateMipmap(GL_TEXTURE_2D);
|
||||
|
||||
|
||||
delete pixel_array;
|
||||
|
||||
|
||||
|
||||
// ========================================= Setup the buffers ==================================================
|
||||
|
||||
int err = 0;
|
||||
|
||||
cl_mem frontWriteBuffer = clCreateFromGLTexture(context , CL_MEM_WRITE_ONLY, GL_TEXTURE_2D, 0, texture, &err);
|
||||
cl_mem frontReadBuffer = clCreateFromGLTexture(context, CL_MEM_READ_ONLY, GL_TEXTURE_2D, 0, texture, &err);
|
||||
cl_mem backBuffer = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(int), (void*)node_grid, &err);
|
||||
|
||||
cl_mem workerCountBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(int), &WORKER_SIZE, &err);
|
||||
cl_mem gridWidthBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(int), &GRID_WIDTH, &err);
|
||||
cl_mem gridHeightBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(int), &GRID_HEIGHT, &err);
|
||||
|
||||
// Compute kernel args
|
||||
status = clSetKernelArg(compute_kernel, 0, sizeof(cl_mem), (void *)&frontWriteBuffer);
|
||||
status = clSetKernelArg(compute_kernel, 1, sizeof(cl_mem), (void *)&backBuffer);
|
||||
status = clSetKernelArg(compute_kernel, 2, sizeof(cl_mem), (void *)&workerCountBuffer);
|
||||
status = clSetKernelArg(compute_kernel, 3, sizeof(cl_mem), (void *)&gridWidthBuffer);
|
||||
status = clSetKernelArg(compute_kernel, 4, sizeof(cl_mem), (void *)&gridHeightBuffer);
|
||||
|
||||
status = clSetKernelArg(align_kernel, 0, sizeof(cl_mem), (void *)&frontReadBuffer);
|
||||
status = clSetKernelArg(align_kernel, 1, sizeof(cl_mem), (void *)&backBuffer);
|
||||
status = clSetKernelArg(align_kernel, 2, sizeof(cl_mem), (void *)&workerCountBuffer);
|
||||
status = clSetKernelArg(align_kernel, 3, sizeof(cl_mem), (void *)&gridWidthBuffer);
|
||||
status = clSetKernelArg(align_kernel, 4, sizeof(cl_mem), (void *)&gridHeightBuffer);
|
||||
|
||||
|
||||
// ===================================== Loop ==================================================================
|
||||
|
||||
while (!glfwWindowShouldClose(gl_window)) {
|
||||
|
||||
glClearColor(0.2f, 0.3f, 0.3f, 1.0f);
|
||||
glClear(GL_COLOR_BUFFER_BIT);
|
||||
|
||||
// ======================================= OpenCL Shtuff ===================================================
|
||||
|
||||
// Work size, for each y line
|
||||
size_t global_work_size[1] = { WORKER_SIZE };
|
||||
|
||||
status = clEnqueueAcquireGLObjects(commandQueue, 1, &frontReadBuffer, 0, 0, 0);
|
||||
status = clEnqueueAcquireGLObjects(commandQueue, 1, &frontWriteBuffer, 0, 0, 0);
|
||||
|
||||
status = clEnqueueNDRangeKernel(commandQueue, compute_kernel, 1, NULL, global_work_size, NULL, 0, NULL, NULL);
|
||||
status = clEnqueueNDRangeKernel(commandQueue, align_kernel, 1, NULL, global_work_size, NULL, 0, NULL, NULL);
|
||||
|
||||
status = clEnqueueReleaseGLObjects(commandQueue, 1, &frontReadBuffer, 0, NULL, NULL);
|
||||
status = clEnqueueReleaseGLObjects(commandQueue, 1, &frontWriteBuffer, 0, NULL, NULL);
|
||||
|
||||
clFinish(commandQueue);
|
||||
|
||||
// ======================================= Rendering Shtuff =================================================
|
||||
|
||||
glfwPollEvents();
|
||||
|
||||
glActiveTexture(GL_TEXTURE0);
|
||||
glBindTexture(GL_TEXTURE_2D, texture);
|
||||
glUniform1i(glGetUniformLocation(vert_frag_shaders.Program, "pixel_texture"), 0);
|
||||
|
||||
// Draw the triangle
|
||||
vert_frag_shaders.Use();
|
||||
glBindVertexArray(VAO);
|
||||
glDrawElements(GL_TRIANGLES, 6, GL_UNSIGNED_INT, 0);
|
||||
glBindVertexArray(0);
|
||||
|
||||
glFinish();
|
||||
|
||||
// Render
|
||||
glfwSwapBuffers(gl_window);
|
||||
|
||||
|
||||
|
||||
}
|
||||
|
||||
glfwTerminate();
|
||||
|
||||
// Release the buffers
|
||||
status = clReleaseMemObject(frontReadBuffer);
|
||||
status = clReleaseMemObject(workerCountBuffer);
|
||||
status = clReleaseMemObject(gridWidthBuffer);
|
||||
status = clReleaseMemObject(gridHeightBuffer);
|
||||
|
||||
// And the program stuff
|
||||
status = clReleaseKernel(compute_kernel);
|
||||
status = clReleaseProgram(compute_program);
|
||||
status = clReleaseProgram(align_program);
|
||||
status = clReleaseCommandQueue(commandQueue);
|
||||
status = clReleaseContext(context);
|
||||
|
||||
if (devices != NULL)
|
||||
{
|
||||
delete devices;
|
||||
devices = NULL;
|
||||
}
|
||||
|
||||
return SUCCESS;
|
||||
}
|
@ -1,90 +0,0 @@
|
||||
#ifndef SHADER_H
|
||||
#define SHADER_H
|
||||
|
||||
#include <string>
|
||||
#include <fstream>
|
||||
#include <sstream>
|
||||
#include <iostream>
|
||||
|
||||
#include <GL/glew.h>
|
||||
|
||||
class Shader {
|
||||
public:
|
||||
GLuint Program;
|
||||
// Constructor generates the shader on the fly
|
||||
Shader(const GLchar* vertexPath, const GLchar* fragmentPath) {
|
||||
// 1. Retrieve the vertex/fragment source code from filePath
|
||||
std::string vertexCode;
|
||||
std::string fragmentCode;
|
||||
std::ifstream vShaderFile;
|
||||
std::ifstream fShaderFile;
|
||||
// ensures ifstream objects can throw exceptions:
|
||||
vShaderFile.exceptions(std::ifstream::badbit);
|
||||
fShaderFile.exceptions(std::ifstream::badbit);
|
||||
try {
|
||||
// Open files
|
||||
vShaderFile.open(vertexPath);
|
||||
fShaderFile.open(fragmentPath);
|
||||
std::stringstream vShaderStream, fShaderStream;
|
||||
// Read file's buffer contents into streams
|
||||
vShaderStream << vShaderFile.rdbuf();
|
||||
fShaderStream << fShaderFile.rdbuf();
|
||||
// close file handlers
|
||||
vShaderFile.close();
|
||||
fShaderFile.close();
|
||||
// Convert stream into string
|
||||
vertexCode = vShaderStream.str();
|
||||
fragmentCode = fShaderStream.str();
|
||||
}
|
||||
catch (std::ifstream::failure e) {
|
||||
std::cout << "ERROR::SHADER::FILE_NOT_SUCCESFULLY_READ" << std::endl;
|
||||
}
|
||||
const GLchar* vShaderCode = vertexCode.c_str();
|
||||
const GLchar * fShaderCode = fragmentCode.c_str();
|
||||
|
||||
GLint success = 0;
|
||||
GLchar infoLog[512];
|
||||
|
||||
// Vertex Shader
|
||||
GLuint vertex = glCreateShader(GL_VERTEX_SHADER);
|
||||
glShaderSource(vertex, 1, &vShaderCode, NULL);
|
||||
glCompileShader(vertex);
|
||||
// Print compile errors if any
|
||||
glGetShaderiv(vertex, GL_COMPILE_STATUS, &success);
|
||||
if (!success) {
|
||||
glGetShaderInfoLog(vertex, 512, NULL, infoLog);
|
||||
std::cout << "ERROR::SHADER::VERTEX::COMPILATION_FAILED\n" << infoLog << std::endl;
|
||||
}
|
||||
// Fragment Shader
|
||||
GLuint fragment = glCreateShader(GL_FRAGMENT_SHADER);
|
||||
glShaderSource(fragment, 1, &fShaderCode, NULL);
|
||||
glCompileShader(fragment);
|
||||
// Print compile errors if any
|
||||
glGetShaderiv(fragment, GL_COMPILE_STATUS, &success);
|
||||
if (!success) {
|
||||
glGetShaderInfoLog(fragment, 512, NULL, infoLog);
|
||||
std::cout << "ERROR::SHADER::FRAGMENT::COMPILATION_FAILED\n" << infoLog << std::endl;
|
||||
}
|
||||
// Shader Program
|
||||
this->Program = glCreateProgram();
|
||||
glAttachShader(this->Program, vertex);
|
||||
glAttachShader(this->Program, fragment);
|
||||
glLinkProgram(this->Program);
|
||||
// Print linking errors if any
|
||||
glGetProgramiv(this->Program, GL_LINK_STATUS, &success);
|
||||
if (!success) {
|
||||
glGetProgramInfoLog(this->Program, 512, NULL, infoLog);
|
||||
std::cout << "ERROR::SHADER::PROGRAM::LINKING_FAILED\n" << infoLog << std::endl;
|
||||
}
|
||||
// Delete the shaders as they're linked into our program now and no longer necessery
|
||||
glDeleteShader(vertex);
|
||||
glDeleteShader(fragment);
|
||||
|
||||
}
|
||||
// Uses the current shader
|
||||
void Use() {
|
||||
glUseProgram(this->Program);
|
||||
}
|
||||
};
|
||||
|
||||
#endif
|
@ -1,20 +0,0 @@
|
||||
__kernel void conway_align(__read_only image2d_t front_image, __global char* back_image, __global int* num_workers, __global int* grid_width, __global int *grid_height)
|
||||
{
|
||||
const sampler_t sampler=CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP | CLK_FILTER_NEAREST;
|
||||
|
||||
// Caclulate the start and end range that this worker will be calculating
|
||||
int data_length = *grid_width * *grid_height;
|
||||
|
||||
int start_range = (data_length / *num_workers) * get_global_id(0);
|
||||
int end_range = (data_length / *num_workers) * (get_global_id(0) + 1);
|
||||
|
||||
|
||||
for (int i = start_range; i < end_range; i++){
|
||||
|
||||
uint4 pixel;
|
||||
pixel = read_imageui(front_image, sampler, (int2)(i,get_global_id(0)));
|
||||
|
||||
back_image[i] = pixel.w / 255;
|
||||
}
|
||||
|
||||
}
|
@ -1,59 +0,0 @@
|
||||
__kernel void conway_compute(__write_only image2d_t front_image, __global char* back_image, __global int* num_workers, __global int* grid_width, __global int *grid_height)
|
||||
{
|
||||
|
||||
float4 black = (float4)(.49, .68, .81, 1);
|
||||
float4 white = (float4)(.49, .68, .71, .3);
|
||||
|
||||
// Caclulate the start and end range that this worker will be calculating
|
||||
int data_length = *grid_width * *grid_height;
|
||||
|
||||
int start_range = (data_length / *num_workers) * get_global_id(0);
|
||||
int end_range = (data_length / *num_workers) * (get_global_id(0) + 1);
|
||||
|
||||
// x, y + 1
|
||||
|
||||
int neighbors = 0;
|
||||
|
||||
for (int i = start_range; i < end_range; i++){
|
||||
|
||||
int2 pixelcoord = (int2) (i % *grid_width, i / *grid_height);
|
||||
|
||||
// add all 8 blocks to neghbors
|
||||
neighbors = 0;
|
||||
// Top
|
||||
neighbors += back_image[i - *grid_width];
|
||||
|
||||
// Top right
|
||||
neighbors += back_image[i - *grid_width + 1];
|
||||
|
||||
// Right
|
||||
neighbors += back_image[i + 1];
|
||||
|
||||
// Bottom Right
|
||||
neighbors += back_image[i + *grid_width + 1];
|
||||
|
||||
// Bottom
|
||||
neighbors += back_image[i + *grid_width];
|
||||
|
||||
// Bottom Left
|
||||
neighbors += back_image[i + *grid_width - 1];
|
||||
|
||||
// Left
|
||||
neighbors += back_image[i - 1];
|
||||
|
||||
// Top left
|
||||
neighbors += back_image[i - *grid_width - 1];
|
||||
|
||||
// push living status to the padded second char
|
||||
|
||||
|
||||
write_imagef(front_image, pixelcoord, black);
|
||||
|
||||
if (neighbors == 3 || (neighbors == 2 && back_image[i])){
|
||||
write_imagef(front_image, pixelcoord, white);
|
||||
}
|
||||
|
||||
//else
|
||||
//write_imagei(front_image, pixelcoord, white);
|
||||
}
|
||||
}
|
@ -1,15 +0,0 @@
|
||||
#version 330 core
|
||||
in vec3 ourColor;
|
||||
in vec2 TexCoord;
|
||||
|
||||
out vec4 color;
|
||||
|
||||
// Texture samplers
|
||||
uniform sampler2D pixel_texture;
|
||||
|
||||
void main()
|
||||
{
|
||||
// Linearly interpolate between both textures (second texture is only slightly combined)
|
||||
//color = vec4(1.0f, 0.5f, 0.2f, 1.0f);
|
||||
color = texture(pixel_texture, TexCoord);
|
||||
}
|
@ -1,16 +0,0 @@
|
||||
#version 330 core
|
||||
layout (location = 0) in vec3 position;
|
||||
layout (location = 1) in vec3 color;
|
||||
layout (location = 2) in vec2 texCoord;
|
||||
|
||||
out vec3 ourColor;
|
||||
out vec2 TexCoord;
|
||||
|
||||
void main()
|
||||
{
|
||||
gl_Position = vec4(position, 1.0f);
|
||||
ourColor = color;
|
||||
// We swap the y-axis by substracing our coordinates from 1. This is done because most images have the top y-axis inversed with OpenGL's top y-axis.
|
||||
// TexCoord = texCoord;
|
||||
TexCoord = vec2(texCoord.x, 1.0 - texCoord.y);
|
||||
}
|