Bart/bullet3
1
0
Fork 0
Code Issues Pull Requests Releases Wiki Activity

Move btOclUtils.cpp/h and btOclCommon.cpp/h to btOpenCLUtils.cpp/h

Browse Source 
Rename member numerator to m_numerator in btConvexHullComputer.cpp to avoid confusion (Thanks to Ian for the feedback)
Add btSoftBody::addAeroForceToNode and btSoftBody::addAeroForceToFace, thanks to Dongsoo Han/Saggita
This commit is contained in:
erwin.coumans
2011-11-14 22:22:49 +00:00
parent a374f8c0d4
commit 0d1fcf7c48
27 changed files with 877 additions and 723 deletions
Download Patch File Download Diff File Expand all files Collapse all files

7
Demos/OpenCLClothDemo/AMD/CMakeLists.txt
View File

@@ -38,10 +38,9 @@ IF (USE_GLUT)
ADD_EXECUTABLE(AppOpenCLClothDemo_AMD
../cl_cloth_demo.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
# ${BULLET_PHYSICS_SOURCE_DIR}/Demos/OpenGL/GLDebugDrawer.cpp
../gl_win.cpp
../clstuff.cpp

7
Demos/OpenCLClothDemo/AMD/premake4.lua
View File

@@ -48,10 +48,9 @@
files {
"../cl_cloth_demo.cpp",
"../../SharedOpenCL/btOclUtils.h",
"../../SharedOpenCL/btOclCommon.h",
"../../SharedOpenCL/btOclUtils.cpp",
"../../SharedOpenCL/btOclCommon.cpp",
"../../SharedOpenCL/btOpenCLUtils.cpp",
"../../SharedOpenCL/btOpenCLUtils.h",
"../../SharedOpenCL/btOpenCLInclude.h",
"../../OpenGL/GLDebugDrawer.cpp",
"../../OpenGL/stb_image.cpp",
"../../OpenGL/stb_image.h",

7
Demos/OpenCLClothDemo/Apple/CMakeLists.txt
View File

@@ -33,10 +33,9 @@ IF (USE_GLUT)
ADD_EXECUTABLE(AppOpenCLClothDemo_Apple
../cl_cloth_demo.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
../gl_win.cpp
../clstuff.cpp
../clstuff.h

7
Demos/OpenCLClothDemo/Intel/CMakeLists.txt
View File

@@ -38,10 +38,9 @@ IF (USE_GLUT)
ADD_EXECUTABLE(AppOpenCLClothDemo_Intel
../cl_cloth_demo.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
# ${BULLET_PHYSICS_SOURCE_DIR}/Demos/OpenGL/GLDebugDrawer.cpp
../gl_win.cpp
../clstuff.cpp

7
Demos/OpenCLClothDemo/Intel/premake4.lua
View File

@@ -48,10 +48,9 @@
files {
"../cl_cloth_demo.cpp",
"../../SharedOpenCL/btOclUtils.h",
"../../SharedOpenCL/btOclCommon.h",
"../../SharedOpenCL/btOclUtils.cpp",
"../../SharedOpenCL/btOclCommon.cpp",
"../../SharedOpenCL/btOpenCLUtils.cpp",
"../../SharedOpenCL/btOpenCLUtils.h",
"../../SharedOpenCL/btOpenCLInclude.h",
"../../OpenGL/GLDebugDrawer.cpp",
"../../OpenGL/stb_image.cpp",
"../../OpenGL/stb_image.h",

7
Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt
View File

@@ -42,10 +42,9 @@ IF (USE_GLUT)
../clstuff.h
../gl_win.h
${BULLET_PHYSICS_SOURCE_DIR}/src/BulletMultiThreaded/GpuSoftBodySolvers/OpenCL/MiniCL/MiniCLTaskWrap.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
)
ELSE (USE_GLUT)

7
Demos/OpenCLClothDemo/NVidia/CMakeLists.txt
View File

@@ -40,10 +40,9 @@ IF (USE_GLUT)
ADD_EXECUTABLE(AppOpenCLClothDemo_NVidia
../cl_cloth_demo.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
../gl_win.cpp
../clstuff.cpp
../clstuff.h

7
Demos/OpenCLClothDemo/NVidia/premake4.lua
View File

@@ -48,10 +48,9 @@
files {
"../cl_cloth_demo.cpp",
"../../SharedOpenCL/btOclUtils.h",
"../../SharedOpenCL/btOclCommon.h",
"../../SharedOpenCL/btOclUtils.cpp",
"../../SharedOpenCL/btOclCommon.cpp",
"../../SharedOpenCL/btOpenCLUtils.cpp",
"../../SharedOpenCL/btOpenCLUtils.h",
"../../SharedOpenCL/btOpenCLInclude.h",
"../../OpenGL/GLDebugDrawer.cpp",
"../../OpenGL/stb_image.cpp",
"../../OpenGL/stb_image.h",

45
Demos/OpenCLClothDemo/clstuff.cpp
View File

@@ -18,9 +18,8 @@ subject to the following restrictions:
#include "clstuff.h"
#include "gl_win.h"
#include "btOpenCLUtils.h"
#include "btOclCommon.h"
#include "btOclUtils.h"
#include "LinearMath/btScalar.h"
#include <stdio.h>
@@ -48,34 +47,24 @@ void initCL( void* glCtx, void* glDC )
#endif//__APPLE__
#endif
g_cxMainContext = btOclCommon::createContextFromType(deviceType, &ciErrNum, glCtx, glDC);
switch (deviceType)
{
case CL_DEVICE_TYPE_GPU:
printf("createContextFromType(CL_DEVICE_TYPE_GPU)\n");
break;
case CL_DEVICE_TYPE_CPU:
printf("createContextFromType(CL_DEVICE_TYPE_CPU)\n");
break;
case CL_DEVICE_TYPE_ALL:
printf("createContextFromType(CL_DEVICE_TYPE_ALL)\n");
break;
default:
printf("createContextFromType(unknown device type %d\n",(int)deviceType);
};
//#endif
g_cxMainContext = btOpenCLUtils::createContextFromType(deviceType, &ciErrNum, glCtx, glDC);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
g_cdDevice = btOclGetDev(g_cxMainContext,0);
btOclPrintDevInfo(g_cdDevice);
int numDev = btOpenCLUtils::getNumDevices(g_cxMainContext);
if (!numDev)
{
btAssert(0);
exit(0);//this is just a demo, exit now
}
g_cdDevice = btOpenCLUtils::getDevice(g_cxMainContext,0);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
btOpenCLDeviceInfo clInfo;
btOpenCLUtils::getDeviceInfo(g_cdDevice,clInfo);
btOpenCLUtils::printDeviceInfo(g_cdDevice);
// create a command-queue
g_cqCommandQue = clCreateCommandQueue(g_cxMainContext, g_cdDevice, 0, &ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);

7
Demos/ParticlesOpenCL/AMD/CMakeLists.txt
View File

@@ -41,10 +41,9 @@ IF (USE_GLUT)
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/main.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.cpp

7
Demos/ParticlesOpenCL/Apple/CMakeLists.txt
View File

@@ -42,15 +42,14 @@ IF (USE_GLUT)
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesSharedDefs.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesSharedTypes.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/main.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesOCL.cl
)
ELSE (USE_GLUT)

7
Demos/ParticlesOpenCL/Intel/CMakeLists.txt
View File

@@ -41,10 +41,9 @@ IF (USE_GLUT)
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/main.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.cpp

7
Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt
View File

@@ -52,10 +52,9 @@ IF (USE_GLUT)
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/MiniCL/MiniCLTaskWrap.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesOCL.cl
)

7
Demos/ParticlesOpenCL/NVidia/CMakeLists.txt
View File

@@ -45,10 +45,9 @@ IF (USE_GLUT)
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/ParticlesDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/shaders.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/ParticlesOpenCL/main.cpp

32
Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp
View File

@@ -27,8 +27,7 @@ subject to the following restrictions:
#endif //__APPLE__
#include "btOclCommon.h"
#include "btOclUtils.h"
#include "btOpenCLUtils.h"
#include "btBulletDynamicsCommon.h"
#include "BulletCollision/CollisionDispatch/btCollisionDispatcher.h"
@@ -337,18 +336,23 @@ void btParticlesDynamicsWorld::initCLKernels(int argc, char** argv)
if (!m_cxMainContext)
{
// m_cxMainContext = clCreateContextFromType(0, CL_DEVICE_TYPE_ALL, NULL, NULL, &ciErrNum);
#ifdef USE_INTEL_OPENCL
m_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_ALL, &ciErrNum);
#else
m_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_GPU, &ciErrNum);
#endif
oclCHECKERROR(ciErrNum, CL_SUCCESS);
m_cdDevice = btOclGetDev(m_cxMainContext,0);
btOclPrintDevInfo(m_cdDevice);
cl_device_type deviceType = CL_DEVICE_TYPE_ALL;
m_cxMainContext = btOpenCLUtils::createContextFromType(deviceType, &ciErrNum, 0, 0);
int numDev = btOpenCLUtils::getNumDevices(m_cxMainContext);
if (!numDev)
{
btAssert(0);
exit(0);//this is just a demo, exit now
}
m_cdDevice = btOpenCLUtils::getDevice(m_cxMainContext,0);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
btOpenCLDeviceInfo clInfo;
btOpenCLUtils::getDeviceInfo(m_cdDevice,clInfo);
btOpenCLUtils::printDeviceInfo(m_cdDevice);
// create a command-queue
m_cqCommandQue = clCreateCommandQueue(m_cxMainContext, m_cdDevice, 0, &ciErrNum);
@@ -440,7 +444,7 @@ void btParticlesDynamicsWorld::initCLKernels(int argc, char** argv)
char cBuildLog[10240];
// char* cPtx;
// size_t szPtxLength;
clGetProgramBuildInfo(m_cpProgram, btOclGetFirstDev(m_cxMainContext), CL_PROGRAM_BUILD_LOG,
clGetProgramBuildInfo(m_cpProgram, m_cdDevice, CL_PROGRAM_BUILD_LOG,
sizeof(cBuildLog), cBuildLog, NULL );
// oclGetProgBinary(m_cpProgram, oclGetFirstDev(m_cxMainContext), &cPtx, &szPtxLength);
// oclLog(LOGBOTH | CLOSELOG, 0.0, "\n\nLog:\n%s\n\n\n\n\nPtx:\n%s\n\n\n", cBuildLog, cPtx);

24
Demos/SerializeDemo/AMD/CMakeLists.txt
View File

@@ -44,10 +44,9 @@ IF (USE_GLUT)
../SerializeDemo.cpp
../SerializeDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/msvc/bullet.rc
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/clew.c
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/clew.h
@@ -57,12 +56,10 @@ IF (USE_GLUT)
../main.cpp
../SerializeDemo.cpp
../SerializeDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/clew.c
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/clew.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
)
ENDIF()
IF (NOT INTERNAL_CREATE_DISTRIBUTABLE_MSVC_PROJECTFILES)
@@ -108,10 +105,9 @@ ELSE (USE_GLUT)
../SerializeDemo.cpp
../SerializeDemo.h
${BULLET_PHYSICS_SOURCE_DIR}/msvc/bullet.rc
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOclCommon.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.cpp
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLUtils.h
${BULLET_PHYSICS_SOURCE_DIR}/Demos/SharedOpenCL/btOpenCLInclude.h
)
ENDIF (USE_GLUT)

37
Demos/SerializeDemo/main.cpp
View File

@@ -23,15 +23,10 @@ subject to the following restrictions:
#ifdef USE_AMD_OPENCL
#ifdef _DEBUG
bool bDebug = true;
#else
bool bDebug = false;
#endif
#include "btOclCommon.h"
#include "btOclUtils.h"
#include "btOpenCLUtils.h"
#include <LinearMath/btScalar.h>
cl_context g_cxMainContext;
@@ -42,6 +37,9 @@ cl_command_queue g_cqCommandQue;
// Returns true if OpenCL is initialized properly, false otherwise.
bool initCL( void* glCtx, void* glDC )
{
const char* vendorSDK = btOpenCLUtils::getSdkVendorName();
printf("This program was compiled using the %s OpenCL SDK\n",vendorSDK);
int ciErrNum = 0;
#ifdef BT_USE_CLEW
@@ -61,22 +59,18 @@ bool initCL( void* glCtx, void* glDC )
cl_device_type deviceType = CL_DEVICE_TYPE_CPU;
#endif
//g_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_ALL, &ciErrNum);
//g_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_GPU, &ciErrNum);
//g_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_CPU, &ciErrNum);
//try CL_DEVICE_TYPE_DEBUG for sequential, non-threaded execution, when using MiniCL on CPU, it gives a full callstack at the crash in the kernel
//#ifdef USE_MINICL
// g_cxMainContext = btOclCommon::createContextFromType(CL_DEVICE_TYPE_DEBUG, &ciErrNum);
//#else
g_cxMainContext = btOclCommon::createContextFromType(deviceType, &ciErrNum, glCtx, glDC);
//#endif
g_cxMainContext = btOpenCLUtils::createContextFromType(deviceType, &ciErrNum, glCtx, glDC);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
g_cdDevice = btOclGetDev(g_cxMainContext,0);
int numDev = btOpenCLUtils::getNumDevices(g_cxMainContext);
if (!numDev)
return false;
g_cdDevice = btOpenCLUtils::getDevice(g_cxMainContext,0);
if ( bDebug ) {
btOclPrintDevInfo(g_cdDevice);
}
btOpenCLDeviceInfo clInfo;
btOpenCLUtils::getDeviceInfo(g_cdDevice,clInfo);
btOpenCLUtils::printDeviceInfo(g_cdDevice);
// create a command-queue
g_cqCommandQue = clCreateCommandQueue(g_cxMainContext, g_cdDevice, 0, &ciErrNum);
@@ -92,6 +86,7 @@ int main(int argc,char** argv)
{
GLDebugDrawer gDebugDrawer;
#ifdef USE_AMD_OPENCL
bool initialized = initCL(0,0);
btAssert(initialized);
#endif //USE_AMD_OPENCL

121
Demos/SharedOpenCL/btOclCommon.cpp
View File

@@ -1,121 +0,0 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2010 Sony Computer Entertainment Inc.
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
Permission is granted to anyone to use this software for any purpose,
including commercial applications, and to alter it and redistribute it freely,
subject to the following restrictions:
1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
3. This notice may not be removed or altered from any source distribution.
*/
#include <string.h>
#include "btOclCommon.h"
static char* spPlatformVendor =
#if defined(CL_PLATFORM_MINI_CL)
"MiniCL, SCEA";
#elif defined(CL_PLATFORM_INTEL)
"Intel(R) Corporation";
#elif defined(CL_PLATFORM_AMD)
"Advanced Micro Devices, Inc.";
#elif defined(CL_PLATFORM_NVIDIA)
"NVIDIA Corporation";
#else
"Unknown Vendor";
#endif
#ifndef CL_PLATFORM_MINI_CL
#ifdef _WIN32
#include "CL/cl_gl.h"
#endif //_WIN32
#endif
cl_context btOclCommon::createContextFromType(cl_device_type deviceType, cl_int* pErrNum, void* pGLContext, void* pGLDC )
{
cl_uint numPlatforms;
cl_platform_id platform = NULL;
cl_int ciErrNum = clGetPlatformIDs(0, NULL, &numPlatforms);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
if(numPlatforms > 0)
{
cl_platform_id* platforms = new cl_platform_id[numPlatforms];
ciErrNum = clGetPlatformIDs(numPlatforms, platforms, NULL);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
for (unsigned i = 0; i < numPlatforms; ++i)
{
char pbuf[128];
ciErrNum = clGetPlatformInfo( platforms[i],
CL_PLATFORM_VENDOR,
sizeof(pbuf),
pbuf,
NULL);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
platform = platforms[i];
if(!strcmp(pbuf, spPlatformVendor))
{
break;
}
}
delete[] platforms;
}
/*
* If we could find our platform, use it. Otherwise pass a NULL and get whatever the
* implementation thinks we should be using.
*/
cl_context_properties cps[7] =
{
CL_CONTEXT_PLATFORM,
(cl_context_properties)platform,
0,
0,
0,
0,
0
};
#ifndef CL_PLATFORM_MINI_CL
#ifdef _WIN32
#ifndef BT_USE_CLEW
// If we have a gl context then enable interop
if( pGLContext )
{
cps[2] = CL_GL_CONTEXT_KHR;
cps[3] = (cl_context_properties)pGLContext;
cps[4] = CL_WGL_HDC_KHR;
cps[5] = (cl_context_properties)pGLDC;
}
#endif // DONT_USE_CLEW
#endif //_WIN32
#endif //CL_PLATFORM_MINI_CL
/* Use NULL for backward compatibility */
cl_context_properties* cprops = (NULL == platform) ? NULL : cps;
cl_context retContext = clCreateContextFromType(cprops,
deviceType,
NULL,
NULL,
&ciErrNum);
if(pErrNum != NULL) *pErrNum = ciErrNum;
return retContext;
}

284
Demos/SharedOpenCL/btOclUtils.cpp
View File

@@ -1,284 +0,0 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2010 Sony Computer Entertainment Inc.
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
Permission is granted to anyone to use this software for any purpose,
including commercial applications, and to alter it and redistribute it freely,
subject to the following restrictions:
1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
3. This notice may not be removed or altered from any source distribution.
*/
#include <stdlib.h>
#include <stdio.h>
#include <string.h>
#define myprintf printf
#include "btOclUtils.h"
//////////////////////////////////////////////////////////////////////////////
//! Gets the id of the nth device from the context
//!
//! @return the id or -1 when out of range
//! @param cxMainContext OpenCL context
//! @param device_idx index of the device of interest
//////////////////////////////////////////////////////////////////////////////
cl_device_id btOclGetDev(cl_context cxMainContext, unsigned int nr)
{
size_t szParmDataBytes;
cl_device_id* cdDevices;
// get the list of GPU devices associated with context
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, 0, NULL, &szParmDataBytes);
if( szParmDataBytes / sizeof(cl_device_id) < nr ) {
return (cl_device_id)-1;
}
cdDevices = (cl_device_id*) malloc(szParmDataBytes);
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, szParmDataBytes, cdDevices, NULL);
cl_device_id device = cdDevices[nr];
free(cdDevices);
return device;
}
//////////////////////////////////////////////////////////////////////////////
//! Loads a Program file and prepends the cPreamble to the code.
//!
//! @return the source string if succeeded, 0 otherwise
//! @param cFilename program filename
//! @param cPreamble code that is prepended to the loaded file, typically a set of #defines or a header
//! @param szFinalLength returned length of the code string
//////////////////////////////////////////////////////////////////////////////
char* btOclLoadProgSource(const char* cFilename, const char* cPreamble, size_t* szFinalLength)
{
// locals
FILE* pFileStream = NULL;
size_t szSourceLength;
// open the OpenCL source code file
pFileStream = fopen(cFilename, "rb");
if(pFileStream == 0)
{
return NULL;
}
size_t szPreambleLength = strlen(cPreamble);
// get the length of the source code
fseek(pFileStream, 0, SEEK_END);
szSourceLength = ftell(pFileStream);
fseek(pFileStream, 0, SEEK_SET);
// allocate a buffer for the source code string and read it in
char* cSourceString = (char *)malloc(szSourceLength + szPreambleLength + 1);
memcpy(cSourceString, cPreamble, szPreambleLength);
fread((cSourceString) + szPreambleLength, szSourceLength, 1, pFileStream);
// close the file and return the total length of the combined (preamble + source) string
fclose(pFileStream);
if(szFinalLength != 0)
{
*szFinalLength = szSourceLength + szPreambleLength;
}
cSourceString[szSourceLength + szPreambleLength] = '\0';
return cSourceString;
}
//////////////////////////////////////////////////////////////////////////////
//! Gets the id of the first device from the context
//!
//! @return the id
//! @param cxMainContext OpenCL context
//////////////////////////////////////////////////////////////////////////////
cl_device_id btOclGetFirstDev(cl_context cxMainContext)
{
size_t szParmDataBytes;
cl_device_id* cdDevices;
// get the list of GPU devices associated with context
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, 0, NULL, &szParmDataBytes);
cdDevices = (cl_device_id*) malloc(szParmDataBytes);
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, szParmDataBytes, cdDevices, NULL);
cl_device_id first = cdDevices[0];
free(cdDevices);
return first;
}
//////////////////////////////////////////////////////////////////////////////
//! Print info about the device
//!
//! @param device OpenCL id of the device
//////////////////////////////////////////////////////////////////////////////
void btOclPrintDevInfo(cl_device_id device)
{
char device_string[1024];
// CL_DEVICE_NAME
clGetDeviceInfo(device, CL_DEVICE_NAME, sizeof(device_string), &device_string, NULL);
myprintf(" CL_DEVICE_NAME: \t\t\t%s\n", device_string);
// CL_DEVICE_VENDOR
clGetDeviceInfo(device, CL_DEVICE_VENDOR, sizeof(device_string), &device_string, NULL);
myprintf(" CL_DEVICE_VENDOR: \t\t\t%s\n", device_string);
// CL_DRIVER_VERSION
clGetDeviceInfo(device, CL_DRIVER_VERSION, sizeof(device_string), &device_string, NULL);
myprintf(" CL_DRIVER_VERSION: \t\t\t%s\n", device_string);
// CL_DEVICE_INFO
cl_device_type type;
clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(type), &type, NULL);
if( type & CL_DEVICE_TYPE_CPU )
myprintf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_CPU");
if( type & CL_DEVICE_TYPE_GPU )
myprintf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_GPU");
if( type & CL_DEVICE_TYPE_ACCELERATOR )
myprintf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_ACCELERATOR");
if( type & CL_DEVICE_TYPE_DEFAULT )
myprintf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_DEFAULT");
// CL_DEVICE_MAX_COMPUTE_UNITS
cl_uint compute_units;
clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(compute_units), &compute_units, NULL);
myprintf(" CL_DEVICE_MAX_COMPUTE_UNITS:\t\t%u\n", compute_units);
// CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS
size_t workitem_dims;
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(workitem_dims), &workitem_dims, NULL);
myprintf(" CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS:\t%u\n", workitem_dims);
// CL_DEVICE_MAX_WORK_ITEM_SIZES
size_t workitem_size[3];
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(workitem_size), &workitem_size, NULL);
myprintf(" CL_DEVICE_MAX_WORK_ITEM_SIZES:\t%u / %u / %u \n", workitem_size[0], workitem_size[1], workitem_size[2]);
// CL_DEVICE_MAX_WORK_GROUP_SIZE
size_t workgroup_size;
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(workgroup_size), &workgroup_size, NULL);
myprintf(" CL_DEVICE_MAX_WORK_GROUP_SIZE:\t%u\n", workgroup_size);
// CL_DEVICE_MAX_CLOCK_FREQUENCY
cl_uint clock_frequency;
clGetDeviceInfo(device, CL_DEVICE_MAX_CLOCK_FREQUENCY, sizeof(clock_frequency), &clock_frequency, NULL);
myprintf(" CL_DEVICE_MAX_CLOCK_FREQUENCY:\t%u MHz\n", clock_frequency);
// CL_DEVICE_ADDRESS_BITS
cl_uint addr_bits;
clGetDeviceInfo(device, CL_DEVICE_ADDRESS_BITS, sizeof(addr_bits), &addr_bits, NULL);
myprintf(" CL_DEVICE_ADDRESS_BITS:\t\t%u\n", addr_bits);
// CL_DEVICE_MAX_MEM_ALLOC_SIZE
cl_ulong max_mem_alloc_size;
clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(max_mem_alloc_size), &max_mem_alloc_size, NULL);
myprintf(" CL_DEVICE_MAX_MEM_ALLOC_SIZE:\t\t%u MByte\n", (unsigned int)(max_mem_alloc_size / (1024 * 1024)));
// CL_DEVICE_GLOBAL_MEM_SIZE
cl_ulong mem_size;
clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(mem_size), &mem_size, NULL);
myprintf(" CL_DEVICE_GLOBAL_MEM_SIZE:\t\t%u MByte\n", (unsigned int)(mem_size / (1024 * 1024)));
// CL_DEVICE_ERROR_CORRECTION_SUPPORT
cl_bool error_correction_support;
clGetDeviceInfo(device, CL_DEVICE_ERROR_CORRECTION_SUPPORT, sizeof(error_correction_support), &error_correction_support, NULL);
myprintf(" CL_DEVICE_ERROR_CORRECTION_SUPPORT:\t%s\n", error_correction_support == CL_TRUE ? "yes" : "no");
// CL_DEVICE_LOCAL_MEM_TYPE
cl_device_local_mem_type local_mem_type;
clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_TYPE, sizeof(local_mem_type), &local_mem_type, NULL);
myprintf(" CL_DEVICE_LOCAL_MEM_TYPE:\t\t%s\n", local_mem_type == 1 ? "local" : "global");
// CL_DEVICE_LOCAL_MEM_SIZE
clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_SIZE, sizeof(mem_size), &mem_size, NULL);
myprintf(" CL_DEVICE_LOCAL_MEM_SIZE:\t\t%u KByte\n", (unsigned int)(mem_size / 1024));
// CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE
clGetDeviceInfo(device, CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE, sizeof(mem_size), &mem_size, NULL);
myprintf(" CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE:\t%u KByte\n", (unsigned int)(mem_size / 1024));
// CL_DEVICE_QUEUE_PROPERTIES
cl_command_queue_properties queue_properties;
clGetDeviceInfo(device, CL_DEVICE_QUEUE_PROPERTIES, sizeof(queue_properties), &queue_properties, NULL);
if( queue_properties & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE )
myprintf(" CL_DEVICE_QUEUE_PROPERTIES:\t\t%s\n", "CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE");
if( queue_properties & CL_QUEUE_PROFILING_ENABLE )
myprintf(" CL_DEVICE_QUEUE_PROPERTIES:\t\t%s\n", "CL_QUEUE_PROFILING_ENABLE");
// CL_DEVICE_IMAGE_SUPPORT
cl_bool image_support;
clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof(image_support), &image_support, NULL);
myprintf(" CL_DEVICE_IMAGE_SUPPORT:\t\t%u\n", image_support);
// CL_DEVICE_MAX_READ_IMAGE_ARGS
cl_uint max_read_image_args;
clGetDeviceInfo(device, CL_DEVICE_MAX_READ_IMAGE_ARGS, sizeof(max_read_image_args), &max_read_image_args, NULL);
myprintf(" CL_DEVICE_MAX_READ_IMAGE_ARGS:\t%u\n", max_read_image_args);
// CL_DEVICE_MAX_WRITE_IMAGE_ARGS
cl_uint max_write_image_args;
clGetDeviceInfo(device, CL_DEVICE_MAX_WRITE_IMAGE_ARGS, sizeof(max_write_image_args), &max_write_image_args, NULL);
myprintf(" CL_DEVICE_MAX_WRITE_IMAGE_ARGS:\t%u\n", max_write_image_args);
// CL_DEVICE_IMAGE2D_MAX_WIDTH, CL_DEVICE_IMAGE2D_MAX_HEIGHT, CL_DEVICE_IMAGE3D_MAX_WIDTH, CL_DEVICE_IMAGE3D_MAX_HEIGHT, CL_DEVICE_IMAGE3D_MAX_DEPTH
size_t szMaxDims[5];
myprintf("\n CL_DEVICE_IMAGE <dim>\n");
clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_WIDTH, sizeof(size_t), &szMaxDims[0], NULL);
myprintf("\t\t\t2D_MAX_WIDTH\t %u\n", szMaxDims[0]);
clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_HEIGHT, sizeof(size_t), &szMaxDims[1], NULL);
myprintf("\t\t\t\t\t2D_MAX_HEIGHT\t %u\n", szMaxDims[1]);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_WIDTH, sizeof(size_t), &szMaxDims[2], NULL);
myprintf("\t\t\t\t\t3D_MAX_WIDTH\t %u\n", szMaxDims[2]);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_HEIGHT, sizeof(size_t), &szMaxDims[3], NULL);
myprintf("\t\t\t\t\t3D_MAX_HEIGHT\t %u\n", szMaxDims[3]);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_DEPTH, sizeof(size_t), &szMaxDims[4], NULL);
myprintf("\t\t\t\t\t3D_MAX_DEPTH\t %u\n", szMaxDims[4]);
// CL_DEVICE_EXTENSIONS: get device extensions, and if any then parse & log the string onto separate lines
clGetDeviceInfo(device, CL_DEVICE_EXTENSIONS, sizeof(device_string), &device_string, NULL);
if (device_string != 0)
{
myprintf("\n CL_DEVICE_EXTENSIONS:%s\n",device_string);
}
else
{
myprintf(" CL_DEVICE_EXTENSIONS: None\n");
}
// CL_DEVICE_PREFERRED_VECTOR_WIDTH_<type>
myprintf(" CL_DEVICE_PREFERRED_VECTOR_WIDTH_<t>\t\n");
cl_uint vec_width [6];
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR, sizeof(cl_uint), &vec_width[0], NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT, sizeof(cl_uint), &vec_width[1], NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT, sizeof(cl_uint), &vec_width[2], NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG, sizeof(cl_uint), &vec_width[3], NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT, sizeof(cl_uint), &vec_width[4], NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE, sizeof(cl_uint), &vec_width[5], NULL);
myprintf("CHAR %u, SHORT %u, INT %u, FLOAT %u, DOUBLE %u\n\n",
vec_width[0], vec_width[1], vec_width[2], vec_width[3], vec_width[4]);
}

42
Demos/SharedOpenCL/btOclUtils.h
View File

@@ -1,42 +0,0 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2010 Sony Computer Entertainment Inc.
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
Permission is granted to anyone to use this software for any purpose,
including commercial applications, and to alter it and redistribute it freely,
subject to the following restrictions:
1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
3. This notice may not be removed or altered from any source distribution.
*/
#ifndef BT_OCL_UTILS_H
#define BT_OCL_UTILS_H
#ifdef USE_MINICL
#include <MiniCL/cl.h>
#else //USE_MINICL
#ifdef BT_USE_CLEW
#include "clew.h"
#else
#ifdef __APPLE__
#include <OpenCL/cl.h>
#else
#include <CL/cl.h>
#endif //__APPLE__
#endif //BT_USE_CLEW
#endif //USE_MINICL
//#define oclCHECKERROR(a, b) btAssert((a) == (b))
#define oclCHECKERROR(a, b) if((a)!=(b)) { printf("OCL Error : %d\n", (a)); btAssert((a) == (b)); }
void btOclPrintDevInfo(cl_device_id device);
cl_device_id btOclGetDev(cl_context cxMainContext, unsigned int nr);
char* btOclLoadProgSource(const char* cFilename, const char* cPreamble, size_t* szFinalLength);
cl_device_id btOclGetFirstDev(cl_context cxMainContext);
#endif //BT_OCL_UTILS_H

44
Demos/SharedOpenCL/btOclCommon.h → Demos/SharedOpenCL/btOpenCLInclude.h
View File

@@ -1,6 +1,6 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2010 Sony Computer Entertainment Inc.
Bullet Continuous Collision Detection and Physics Library
Copyright (c) 2011 Advanced Micro Devices, Inc. http://bulletphysics.org
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
@@ -13,21 +13,31 @@ subject to the following restrictions:
3. This notice may not be removed or altered from any source distribution.
*/
#ifndef BTOCLCOMMON_H
#define BTOCLCOMMON_H
#include "btOclUtils.h"
class btOclCommon
{
public:
// CL Context optionally takes a GL context. This is a generic type because we don't really want this code
// to have to understand GL types.
// It is a HGLRC in _WIN32 or a GLXContext otherwise.
static cl_context createContextFromType(cl_device_type deviceType, cl_int* pErrNum, void* pGLCtx = 0, void* pGLDC = 0);
};
#ifndef BT_OPENCL_INCLUDE_H
#define BT_OPENCL_INCLUDE_H
#ifdef __APPLE__
#ifdef USE_MINICL
#include <MiniCL/cl.h>
#else
#include <OpenCL/cl.h>
#endif
#else
#ifdef USE_MINICL
#include <MiniCL/cl.h>
#else
#include <CL/cl.h>
#ifdef _WIN32
#include "CL/cl_gl.h"
#endif //_WIN32
#endif
#endif //__APPLE__
#include <assert.h>
#include <stdio.h>
#define oclCHECKERROR(a, b) if((a)!=(b)) { printf("OCL Error : %d\n", (a)); assert((a) == (b)); }
#endif //BT_OPENCL_INCLUDE_H
#endif // BTOCLCOMMON_H

451
Demos/SharedOpenCL/btOpenCLUtils.cpp Normal file
View File

@@ -0,0 +1,451 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2011 Sony Computer Entertainment Inc.
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
Permission is granted to anyone to use this software for any purpose,
including commercial applications, and to alter it and redistribute it freely,
subject to the following restrictions:
1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
3. This notice may not be removed or altered from any source distribution.
*/
//original author: Roman Ponomarev
//cleanup by Erwin Coumans
#include <string.h>
#include "btOpenCLUtils.h"
#include <stdio.h>
#include <stdlib.h>
//Set the preferred platform vendor using the OpenCL SDK
static char* spPlatformVendor =
#if defined(CL_PLATFORM_MINI_CL)
"MiniCL, SCEA";
#elif defined(CL_PLATFORM_AMD)
"Advanced Micro Devices, Inc.";
#elif defined(CL_PLATFORM_NVIDIA)
"NVIDIA Corporation";
#elif defined(CL_PLATFORM_INTEL)
"Intel(R) Corporation";
#else
"Unknown Vendor";
#endif
#ifndef CL_PLATFORM_MINI_CL
#ifdef _WIN32
#include "CL/cl_gl.h"
#endif //_WIN32
#endif
int btOpenCLUtils::getNumPlatforms(cl_int* pErrNum)
{
cl_uint numPlatforms=0;
cl_int ciErrNum = clGetPlatformIDs(0, NULL, &numPlatforms);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL)
*pErrNum = ciErrNum;
}
return numPlatforms;
}
const char* btOpenCLUtils::getSdkVendorName()
{
return spPlatformVendor;
}
cl_platform_id btOpenCLUtils::getPlatform(int platformIndex, cl_int* pErrNum)
{
cl_platform_id platform = 0;
cl_uint numPlatforms;
cl_int ciErrNum = clGetPlatformIDs(0, NULL, &numPlatforms);
if (platformIndex>=0 && platformIndex<numPlatforms)
{
cl_platform_id* platforms = new cl_platform_id[numPlatforms];
ciErrNum = clGetPlatformIDs(numPlatforms, platforms, NULL);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL)
*pErrNum = ciErrNum;
return platform;
}
platform = platforms[platformIndex];
delete[] platforms;
}
return platform;
}
void btOpenCLUtils::getPlatformInfo(cl_platform_id platform, btOpenCLPlatformInfo& platformInfo)
{
cl_int ciErrNum;
ciErrNum = clGetPlatformInfo( platform,CL_PLATFORM_VENDOR,BT_MAX_STRING_LENGTH,platformInfo.m_platformVendor,NULL);
oclCHECKERROR(ciErrNum,CL_SUCCESS);
ciErrNum = clGetPlatformInfo( platform,CL_PLATFORM_NAME,BT_MAX_STRING_LENGTH,platformInfo.m_platformName,NULL);
oclCHECKERROR(ciErrNum,CL_SUCCESS);
ciErrNum = clGetPlatformInfo( platform,CL_PLATFORM_VERSION,BT_MAX_STRING_LENGTH,platformInfo.m_platformVersion,NULL);
oclCHECKERROR(ciErrNum,CL_SUCCESS);
}
cl_context btOpenCLUtils::createContextFromPlatform(cl_platform_id platform, cl_device_type deviceType, cl_int* pErrNum, void* pGLContext, void* pGLDC)
{
cl_context retContext = 0;
cl_int ciErrNum=0;
/*
* If we could find our platform, use it. Otherwise pass a NULL and get whatever the
* implementation thinks we should be using.
*/
cl_context_properties cps[7] = {0,0,0,0,0,0,0};
#ifndef CL_PLATFORM_MINI_CL
cps[0] = CL_CONTEXT_PLATFORM;
cps[1] = (cl_context_properties)platform;
if (pGLContext && pGLDC)
{
cps[2] = CL_GL_CONTEXT_KHR;
cps[3] = (cl_context_properties)pGLContext;
cps[4] = CL_WGL_HDC_KHR;
cps[5] = (cl_context_properties)pGLDC;
}
#endif //CL_PLATFORM_MINI_CL
cl_context_properties* cprops = (NULL == platform) ? NULL : cps;
retContext = clCreateContextFromType(cprops,
deviceType,
NULL,
NULL,
&ciErrNum);
if(pErrNum != NULL)
{
*pErrNum = ciErrNum;
};
return retContext;
}
cl_context btOpenCLUtils::createContextFromType(cl_device_type deviceType, cl_int* pErrNum, void* pGLContext, void* pGLDC )
{
cl_uint numPlatforms;
cl_context retContext = 0;
cl_int ciErrNum = clGetPlatformIDs(0, NULL, &numPlatforms);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
if(numPlatforms > 0)
{
cl_platform_id* platforms = new cl_platform_id[numPlatforms];
ciErrNum = clGetPlatformIDs(numPlatforms, platforms, NULL);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
int i;
for ( i = 0; i < numPlatforms; ++i)
{
char pbuf[128];
ciErrNum = clGetPlatformInfo( platforms[i],
CL_PLATFORM_VENDOR,
sizeof(pbuf),
pbuf,
NULL);
if(ciErrNum != CL_SUCCESS)
{
if(pErrNum != NULL) *pErrNum = ciErrNum;
return NULL;
}
if(!strcmp(pbuf, spPlatformVendor))
{
cl_platform_id tmpPlatform = platforms[0];
platforms[0] = platforms[i];
platforms[i] = tmpPlatform;
break;
}
}
for (i = 0; i < numPlatforms; ++i)
{
cl_platform_id platform = platforms[i];
assert(platform);
retContext = btOpenCLUtils::createContextFromPlatform(platform,deviceType,pErrNum,pGLContext,pGLDC);
if (retContext)
{
printf("OpenCL platform details:\n");
btOpenCLPlatformInfo platformInfo;
btOpenCLUtils::getPlatformInfo(platform, platformInfo);
printf(" CL_PLATFORM_VENDOR: \t\t\t%s\n",platformInfo.m_platformVendor);
printf(" CL_PLATFORM_NAME: \t\t\t%s\n",platformInfo.m_platformName);
printf(" CL_PLATFORM_VERSION: \t\t\t%s\n",platformInfo.m_platformVersion);
break;
}
}
delete[] platforms;
}
return retContext;
}
//////////////////////////////////////////////////////////////////////////////
//! Gets the id of the nth device from the context
//!
//! @return the id or -1 when out of range
//! @param cxMainContext OpenCL context
//! @param device_idx index of the device of interest
//////////////////////////////////////////////////////////////////////////////
cl_device_id btOpenCLUtils::getDevice(cl_context cxMainContext, int deviceIndex)
{
size_t szParmDataBytes;
cl_device_id* cdDevices;
// get the list of devices associated with context
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, 0, NULL, &szParmDataBytes);
if( szParmDataBytes / sizeof(cl_device_id) < deviceIndex ) {
return (cl_device_id)-1;
}
cdDevices = (cl_device_id*) malloc(szParmDataBytes);
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, szParmDataBytes, cdDevices, NULL);
cl_device_id device = cdDevices[deviceIndex];
free(cdDevices);
return device;
}
int btOpenCLUtils::getNumDevices(cl_context cxMainContext)
{
size_t szParamDataBytes;
clGetContextInfo(cxMainContext, CL_CONTEXT_DEVICES, 0, NULL, &szParamDataBytes);
int device_count = (int) szParamDataBytes/ sizeof(cl_device_id);
return device_count;
}
void btOpenCLUtils::printDeviceInfo(cl_device_id device)
{
btOpenCLDeviceInfo info;
getDeviceInfo(device,info);
printf(" CL_DEVICE_NAME: \t\t\t%s\n", info.m_deviceName);
printf(" CL_DEVICE_VENDOR: \t\t\t%s\n", info.m_deviceVendor);
printf(" CL_DRIVER_VERSION: \t\t\t%s\n", info.m_driverVersion);
if( info.m_deviceType & CL_DEVICE_TYPE_CPU )
printf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_CPU");
if( info.m_deviceType & CL_DEVICE_TYPE_GPU )
printf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_GPU");
if( info.m_deviceType & CL_DEVICE_TYPE_ACCELERATOR )
printf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_ACCELERATOR");
if( info.m_deviceType & CL_DEVICE_TYPE_DEFAULT )
printf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_DEFAULT");
printf(" CL_DEVICE_MAX_COMPUTE_UNITS:\t\t%u\n", info.m_computeUnits);
printf(" CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS:\t%u\n", info.m_workitemDims);
printf(" CL_DEVICE_MAX_WORK_ITEM_SIZES:\t%u / %u / %u \n", info.m_workItemSize[0], info.m_workItemSize[1], info.m_workItemSize[2]);
printf(" CL_DEVICE_MAX_WORK_GROUP_SIZE:\t%u\n", info.m_workgroupSize);
printf(" CL_DEVICE_MAX_CLOCK_FREQUENCY:\t%u MHz\n", info.m_clockFrequency);
printf(" CL_DEVICE_ADDRESS_BITS:\t\t%u\n", info.m_addressBits);
printf(" CL_DEVICE_MAX_MEM_ALLOC_SIZE:\t\t%u MByte\n", (unsigned int)(info.m_maxMemAllocSize/ (1024 * 1024)));
printf(" CL_DEVICE_GLOBAL_MEM_SIZE:\t\t%u MByte\n", (unsigned int)(info.m_globalMemSize/ (1024 * 1024)));
printf(" CL_DEVICE_ERROR_CORRECTION_SUPPORT:\t%s\n", info.m_errorCorrectionSupport== CL_TRUE ? "yes" : "no");
printf(" CL_DEVICE_LOCAL_MEM_TYPE:\t\t%s\n", info.m_localMemType == 1 ? "local" : "global");
printf(" CL_DEVICE_LOCAL_MEM_SIZE:\t\t%u KByte\n", (unsigned int)(info.m_localMemSize / 1024));
printf(" CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE:\t%u KByte\n", (unsigned int)(info.m_constantBufferSize / 1024));
if( info.m_queueProperties & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE )
printf(" CL_DEVICE_QUEUE_PROPERTIES:\t\t%s\n", "CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE");
if( info.m_queueProperties & CL_QUEUE_PROFILING_ENABLE )
printf(" CL_DEVICE_QUEUE_PROPERTIES:\t\t%s\n", "CL_QUEUE_PROFILING_ENABLE");
printf(" CL_DEVICE_IMAGE_SUPPORT:\t\t%u\n", info.m_imageSupport);
printf(" CL_DEVICE_MAX_READ_IMAGE_ARGS:\t%u\n", info.m_maxReadImageArgs);
printf(" CL_DEVICE_MAX_WRITE_IMAGE_ARGS:\t%u\n", info.m_maxWriteImageArgs);
printf("\n CL_DEVICE_IMAGE <dim>");
printf("\t\t\t2D_MAX_WIDTH\t %u\n", info.m_image2dMaxWidth);
printf("\t\t\t\t\t2D_MAX_HEIGHT\t %u\n", info.m_image2dMaxHeight);
printf("\t\t\t\t\t3D_MAX_WIDTH\t %u\n", info.m_image3dMaxWidth);
printf("\t\t\t\t\t3D_MAX_HEIGHT\t %u\n", info.m_image3dMaxHeight);
printf("\t\t\t\t\t3D_MAX_DEPTH\t %u\n", info.m_image3dMaxDepth);
if (info.m_deviceExtensions != 0)
printf("\n CL_DEVICE_EXTENSIONS:%s\n",info.m_deviceExtensions);
else
printf(" CL_DEVICE_EXTENSIONS: None\n");
printf(" CL_DEVICE_PREFERRED_VECTOR_WIDTH_<t>\t");
printf("CHAR %u, SHORT %u, INT %u,LONG %u, FLOAT %u, DOUBLE %u\n\n\n",
info.m_vecWidthChar, info.m_vecWidthShort, info.m_vecWidthInt, info.m_vecWidthLong,info.m_vecWidthFloat, info.m_vecWidthDouble);
}
void btOpenCLUtils::getDeviceInfo(cl_device_id device, btOpenCLDeviceInfo& info)
{
// CL_DEVICE_NAME
clGetDeviceInfo(device, CL_DEVICE_NAME, BT_MAX_STRING_LENGTH, &info.m_deviceName, NULL);
// CL_DEVICE_VENDOR
clGetDeviceInfo(device, CL_DEVICE_VENDOR, BT_MAX_STRING_LENGTH, &info.m_deviceVendor, NULL);
// CL_DRIVER_VERSION
clGetDeviceInfo(device, CL_DRIVER_VERSION, BT_MAX_STRING_LENGTH, &info.m_driverVersion, NULL);
// CL_DEVICE_INFO
clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(cl_device_type), &info.m_deviceType, NULL);
// CL_DEVICE_MAX_COMPUTE_UNITS
clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(info.m_computeUnits), &info.m_computeUnits, NULL);
// CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(info.m_workitemDims), &info.m_workitemDims, NULL);
// CL_DEVICE_MAX_WORK_ITEM_SIZES
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(info.m_workItemSize), &info.m_workItemSize, NULL);
// CL_DEVICE_MAX_WORK_GROUP_SIZE
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(info.m_workgroupSize), &info.m_workgroupSize, NULL);
// CL_DEVICE_MAX_CLOCK_FREQUENCY
clGetDeviceInfo(device, CL_DEVICE_MAX_CLOCK_FREQUENCY, sizeof(info.m_clockFrequency), &info.m_clockFrequency, NULL);
// CL_DEVICE_ADDRESS_BITS
clGetDeviceInfo(device, CL_DEVICE_ADDRESS_BITS, sizeof(info.m_addressBits), &info.m_addressBits, NULL);
// CL_DEVICE_MAX_MEM_ALLOC_SIZE
clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(info.m_maxMemAllocSize), &info.m_maxMemAllocSize, NULL);
// CL_DEVICE_GLOBAL_MEM_SIZE
clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(info.m_globalMemSize), &info.m_globalMemSize, NULL);
// CL_DEVICE_ERROR_CORRECTION_SUPPORT
clGetDeviceInfo(device, CL_DEVICE_ERROR_CORRECTION_SUPPORT, sizeof(info.m_errorCorrectionSupport), &info.m_errorCorrectionSupport, NULL);
// CL_DEVICE_LOCAL_MEM_TYPE
clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_TYPE, sizeof(info.m_localMemType), &info.m_localMemType, NULL);
// CL_DEVICE_LOCAL_MEM_SIZE
clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_SIZE, sizeof(info.m_localMemSize), &info.m_localMemSize, NULL);
// CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE
clGetDeviceInfo(device, CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE, sizeof(info.m_constantBufferSize), &info.m_constantBufferSize, NULL);
// CL_DEVICE_QUEUE_PROPERTIES
clGetDeviceInfo(device, CL_DEVICE_QUEUE_PROPERTIES, sizeof(info.m_queueProperties), &info.m_queueProperties, NULL);
// CL_DEVICE_IMAGE_SUPPORT
clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof(info.m_imageSupport), &info.m_imageSupport, NULL);
// CL_DEVICE_MAX_READ_IMAGE_ARGS
clGetDeviceInfo(device, CL_DEVICE_MAX_READ_IMAGE_ARGS, sizeof(info.m_maxReadImageArgs), &info.m_maxReadImageArgs, NULL);
// CL_DEVICE_MAX_WRITE_IMAGE_ARGS
clGetDeviceInfo(device, CL_DEVICE_MAX_WRITE_IMAGE_ARGS, sizeof(info.m_maxWriteImageArgs), &info.m_maxWriteImageArgs, NULL);
// CL_DEVICE_IMAGE2D_MAX_WIDTH, CL_DEVICE_IMAGE2D_MAX_HEIGHT, CL_DEVICE_IMAGE3D_MAX_WIDTH, CL_DEVICE_IMAGE3D_MAX_HEIGHT, CL_DEVICE_IMAGE3D_MAX_DEPTH
clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_WIDTH, sizeof(size_t), &info.m_image2dMaxWidth, NULL);
clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_HEIGHT, sizeof(size_t), &info.m_image2dMaxHeight, NULL);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_WIDTH, sizeof(size_t), &info.m_image3dMaxWidth, NULL);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_HEIGHT, sizeof(size_t), &info.m_image3dMaxHeight, NULL);
clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_DEPTH, sizeof(size_t), &info.m_image3dMaxDepth, NULL);
// CL_DEVICE_EXTENSIONS: get device extensions, and if any then parse & log the string onto separate lines
clGetDeviceInfo(device, CL_DEVICE_EXTENSIONS, BT_MAX_STRING_LENGTH, &info.m_deviceExtensions, NULL);
// CL_DEVICE_PREFERRED_VECTOR_WIDTH_<type>
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR, sizeof(cl_uint), &info.m_vecWidthChar, NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT, sizeof(cl_uint), &info.m_vecWidthShort, NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT, sizeof(cl_uint), &info.m_vecWidthInt, NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG, sizeof(cl_uint), &info.m_vecWidthLong, NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT, sizeof(cl_uint), &info.m_vecWidthFloat, NULL);
clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE, sizeof(cl_uint), &info.m_vecWidthDouble, NULL);
}
cl_kernel btOpenCLUtils::compileCLKernelFromString(cl_context clContext, const char* kernelSource, const char* kernelName, const char* additionalMacros )
{
printf("compiling kernelName: %s ",kernelName);
cl_kernel kernel;
cl_int ciErrNum;
size_t program_length = strlen(kernelSource);
cl_program m_cpProgram = clCreateProgramWithSource(clContext, 1, (const char**)&kernelSource, &program_length, &ciErrNum);
// oclCHECKERROR(ciErrNum, CL_SUCCESS);
// Build the program with 'mad' Optimization option
#ifdef MAC
char* flags = "-cl-mad-enable -DMAC -DGUID_ARG";
#else
//const char* flags = "-DGUID_ARG= -fno-alias";
const char* flags = "-DGUID_ARG= ";
#endif
char* compileFlags = new char[strlen(additionalMacros) + strlen(flags) + 5];
sprintf(compileFlags, "%s %s", flags, additionalMacros);
ciErrNum = clBuildProgram(m_cpProgram, 0, NULL, compileFlags, NULL, NULL);
if (ciErrNum != CL_SUCCESS)
{
size_t numDevices;
clGetProgramInfo( m_cpProgram, CL_PROGRAM_DEVICES, 0, 0, &numDevices );
cl_device_id *devices = new cl_device_id[numDevices];
clGetProgramInfo( m_cpProgram, CL_PROGRAM_DEVICES, numDevices, devices, &numDevices );
for( int i = 0; i < 2; ++i )
{
char *build_log;
size_t ret_val_size;
clGetProgramBuildInfo(m_cpProgram, devices[i], CL_PROGRAM_BUILD_LOG, 0, NULL, &ret_val_size);
build_log = new char[ret_val_size+1];
clGetProgramBuildInfo(m_cpProgram, devices[i], CL_PROGRAM_BUILD_LOG, ret_val_size, build_log, NULL);
// to be carefully, terminate with \0
// there's no information in the reference whether the string is 0 terminated or not
build_log[ret_val_size] = '\0';
printf("Error in clBuildProgram, Line %u in file %s, Log: \n%s\n !!!\n\n", __LINE__, __FILE__, build_log);
delete[] build_log;
}
assert(0);
exit(0);
}
// Create the kernel
kernel = clCreateKernel(m_cpProgram, kernelName, &ciErrNum);
if (ciErrNum != CL_SUCCESS)
{
printf("Error in clCreateKernel, Line %u in file %s !!!\n\n", __LINE__, __FILE__);
assert(0);
exit(0);
}
clReleaseProgram(m_cpProgram);
printf("ready. \n");
delete [] compileFlags;
return kernel;
}

101
Demos/SharedOpenCL/btOpenCLUtils.h Normal file
View File

@@ -0,0 +1,101 @@
/*
Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
Copyright (C) 2006 - 2011 Sony Computer Entertainment Inc.
This software is provided 'as-is', without any express or implied warranty.
In no event will the authors be held liable for any damages arising from the use of this software.
Permission is granted to anyone to use this software for any purpose,
including commercial applications, and to alter it and redistribute it freely,
subject to the following restrictions:
1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
3. This notice may not be removed or altered from any source distribution.
*/
//original author: Roman Ponomarev
//cleanup by Erwin Coumans
#ifndef BT_OPENCL_UTILS_H
#define BT_OPENCL_UTILS_H
#include "btOpenCLInclude.h"
#define BT_MAX_STRING_LENGTH 1024
struct btOpenCLDeviceInfo
{
char m_deviceName[BT_MAX_STRING_LENGTH];
char m_deviceVendor[BT_MAX_STRING_LENGTH];
char m_driverVersion[BT_MAX_STRING_LENGTH];
char m_deviceExtensions[BT_MAX_STRING_LENGTH];
cl_device_type m_deviceType;
cl_uint m_computeUnits;
size_t m_workitemDims;
size_t m_workItemSize[3];
size_t m_image2dMaxWidth;
size_t m_image2dMaxHeight;
size_t m_image3dMaxWidth;
size_t m_image3dMaxHeight;
size_t m_image3dMaxDepth;
size_t m_workgroupSize;
cl_uint m_clockFrequency;
cl_ulong m_constantBufferSize;
cl_ulong m_localMemSize;
cl_ulong m_globalMemSize;
cl_bool m_errorCorrectionSupport;
cl_device_local_mem_type m_localMemType;
cl_uint m_maxReadImageArgs;
cl_uint m_maxWriteImageArgs;
cl_uint m_addressBits;
cl_ulong m_maxMemAllocSize;
cl_command_queue_properties m_queueProperties;
cl_bool m_imageSupport;
cl_uint m_vecWidthChar;
cl_uint m_vecWidthShort;
cl_uint m_vecWidthInt;
cl_uint m_vecWidthLong;
cl_uint m_vecWidthFloat;
cl_uint m_vecWidthDouble;
};
struct btOpenCLPlatformInfo
{
char m_platformVendor[BT_MAX_STRING_LENGTH];
char m_platformName[BT_MAX_STRING_LENGTH];
char m_platformVersion[BT_MAX_STRING_LENGTH];
};
class btOpenCLUtils
{
public:
/// CL Context optionally takes a GL context. This is a generic type because we don't really want this code
/// to have to understand GL types. It is a HGLRC in _WIN32 or a GLXContext otherwise.
static cl_context createContextFromType(cl_device_type deviceType, cl_int* pErrNum, void* pGLCtx = 0, void* pGLDC = 0);
static int getNumDevices(cl_context cxMainContext);
static cl_device_id getDevice(cl_context cxMainContext, int nr);
static void getDeviceInfo(cl_device_id device, btOpenCLDeviceInfo& info);
static void printDeviceInfo(cl_device_id device);
static cl_kernel compileCLKernelFromString( cl_context clContext,const char* kernelSource, const char* kernelName, const char* additionalMacros = "" );
//the following optional APIs provide access using specific platform information
static int getNumPlatforms(cl_int* pErrNum=0);
///get the nr'th platform, where nr is in the range [0..getNumPlatforms)
static cl_platform_id getPlatform(int nr, cl_int* pErrNum=0);
static void getPlatformInfo(cl_platform_id platform, btOpenCLPlatformInfo& platformInfo);
static const char* getSdkVendorName();
static cl_context createContextFromPlatform(cl_platform_id platform, cl_device_type deviceType, cl_int* pErrNum, void* pGLCtx = 0, void* pGLDC = 0);
};
#endif // BT_OPENCL_UTILS_H

284
src/BulletSoftBody/btSoftBody.cpp
View File

@@ -453,6 +453,167 @@ void btSoftBody::addForce(const btVector3& force,int node)
}
}
void btSoftBody::addAeroForceToNode(const btVector3& windVelocity,int nodeIndex)
{
btAssert(nodeIndex >= 0 && nodeIndex < m_nodes.size());
const btScalar dt = m_sst.sdt;
const btScalar kLF = m_cfg.kLF;
const btScalar kDG = m_cfg.kDG;
const btScalar kPR = m_cfg.kPR;
const btScalar kVC = m_cfg.kVC;
const bool as_lift = kLF>0;
const bool as_drag = kDG>0;
const bool as_aero = as_lift || as_drag;
const bool as_vaero = as_aero && (m_cfg.aeromodel < btSoftBody::eAeroModel::F_TwoSided);
Node& n = m_nodes[nodeIndex];
if( n.m_im>0 )
{
btSoftBody::sMedium medium;
EvaluateMedium(m_worldInfo, n.m_x, medium);
medium.m_velocity = windVelocity;
medium.m_density = m_worldInfo->air_density;
/* Aerodynamics */
if(as_vaero)
{
const btVector3 rel_v = n.m_v - medium.m_velocity;
const btScalar rel_v_len = rel_v.length();
const btScalar rel_v2 = rel_v.length2();
if(rel_v2>SIMD_EPSILON)
{
const btVector3 rel_v_nrm = rel_v.normalized();
btVector3 nrm = n.m_n;
if (m_cfg.aeromodel == btSoftBody::eAeroModel::V_TwoSidedLiftDrag)
{
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
btVector3 fDrag(0, 0, 0);
btVector3 fLift(0, 0, 0);
btScalar n_dot_v = nrm.dot(rel_v_nrm);
btScalar tri_area = 0.5f * n.m_area;
fDrag = 0.5f * kDG * medium.m_density * rel_v2 * tri_area * n_dot_v * (-rel_v_nrm);
// Check angle of attack
// cos(10<31>) = 0.98480
if ( 0 < n_dot_v && n_dot_v < 0.98480f)
fLift = 0.5f * kLF * medium.m_density * rel_v_len * tri_area * btSqrt(1.0f-n_dot_v*n_dot_v) * (nrm.cross(rel_v_nrm).cross(rel_v_nrm));
n.m_f += fDrag;
n.m_f += fLift;
}
else if (m_cfg.aeromodel == btSoftBody::eAeroModel::V_Point || m_cfg.aeromodel == btSoftBody::eAeroModel::V_OneSided || m_cfg.aeromodel == btSoftBody::eAeroModel::V_TwoSided)
{
if (btSoftBody::eAeroModel::V_TwoSided)
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
const btScalar dvn = btDot(rel_v,nrm);
/* Compute forces */
if(dvn>0)
{
btVector3 force(0,0,0);
const btScalar c0 = n.m_area * dvn * rel_v2/2;
const btScalar c1 = c0 * medium.m_density;
force += nrm*(-c1*kLF);
force += rel_v.normalized() * (-c1 * kDG);
ApplyClampedForce(n, force, dt);
}
}
}
}
}
}
void btSoftBody::addAeroForceToFace(const btVector3& windVelocity,int faceIndex)
{
const btScalar dt = m_sst.sdt;
const btScalar kLF = m_cfg.kLF;
const btScalar kDG = m_cfg.kDG;
const btScalar kPR = m_cfg.kPR;
const btScalar kVC = m_cfg.kVC;
const bool as_lift = kLF>0;
const bool as_drag = kDG>0;
const bool as_aero = as_lift || as_drag;
const bool as_faero = as_aero && (m_cfg.aeromodel >= btSoftBody::eAeroModel::F_TwoSided);
if(as_faero)
{
btSoftBody::Face& f=m_faces[faceIndex];
btSoftBody::sMedium medium;
const btVector3 v=(f.m_n[0]->m_v+f.m_n[1]->m_v+f.m_n[2]->m_v)/3;
const btVector3 x=(f.m_n[0]->m_x+f.m_n[1]->m_x+f.m_n[2]->m_x)/3;
EvaluateMedium(m_worldInfo,x,medium);
medium.m_velocity = windVelocity;
medium.m_density = m_worldInfo->air_density;
const btVector3 rel_v=v-medium.m_velocity;
const btScalar rel_v_len = rel_v.length();
const btScalar rel_v2=rel_v.length2();
if(rel_v2>SIMD_EPSILON)
{
const btVector3 rel_v_nrm = rel_v.normalized();
btVector3 nrm = f.m_normal;
if (m_cfg.aeromodel == btSoftBody::eAeroModel::F_TwoSidedLiftDrag)
{
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
btVector3 fDrag(0, 0, 0);
btVector3 fLift(0, 0, 0);
btScalar n_dot_v = nrm.dot(rel_v_nrm);
btScalar tri_area = 0.5f * f.m_ra;
fDrag = 0.5f * kDG * medium.m_density * rel_v2 * tri_area * n_dot_v * (-rel_v_nrm);
// Check angle of attack
// cos(10<31>) = 0.98480
if ( 0 < n_dot_v && n_dot_v < 0.98480f)
fLift = 0.5f * kLF * medium.m_density * rel_v_len * tri_area * btSqrt(1.0f-n_dot_v*n_dot_v) * (nrm.cross(rel_v_nrm).cross(rel_v_nrm));
fDrag /= 3;
fLift /= 3;
for(int j=0;j<3;++j)
{
if (f.m_n[j]->m_im>0)
{
f.m_n[j]->m_f += fDrag;
f.m_n[j]->m_f += fLift;
}
}
}
else if (m_cfg.aeromodel == btSoftBody::eAeroModel::F_OneSided || m_cfg.aeromodel == btSoftBody::eAeroModel::F_TwoSided)
{
if (btSoftBody::eAeroModel::F_TwoSided)
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
const btScalar dvn=btDot(rel_v,nrm);
/* Compute forces */
if(dvn>0)
{
btVector3 force(0,0,0);
const btScalar c0 = f.m_ra*dvn*rel_v2;
const btScalar c1 = c0*medium.m_density;
force += nrm*(-c1*kLF);
force += rel_v.normalized()*(-c1*kDG);
force /= 3;
for(int j=0;j<3;++j) ApplyClampedForce(*f.m_n[j],force,dt);
}
}
}
}
}
//
void btSoftBody::addVelocity(const btVector3& velocity)
{
@@ -2694,60 +2855,8 @@ void btSoftBody::applyForces()
{
if(use_medium)
{
EvaluateMedium(m_worldInfo, n.m_x, medium);
medium.m_velocity = m_windVelocity;
medium.m_density = m_worldInfo->air_density;
/* Aerodynamics */
if(as_vaero)
{
const btVector3 rel_v = n.m_v - medium.m_velocity;
const btScalar rel_v_len = rel_v.length();
const btScalar rel_v2 = rel_v.length2();
if(rel_v2>SIMD_EPSILON)
{
const btVector3 rel_v_nrm = rel_v.normalized();
btVector3 nrm = n.m_n;
if (m_cfg.aeromodel == btSoftBody::eAeroModel::V_TwoSidedLiftDrag)
{
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
btVector3 fDrag(0, 0, 0);
btVector3 fLift(0, 0, 0);
btScalar n_dot_v = nrm.dot(rel_v_nrm);
btScalar tri_area = 0.5f * n.m_area;
fDrag = 0.5f * kDG * medium.m_density * rel_v2 * tri_area * n_dot_v * (-rel_v_nrm);
// Check angle of attack
// cos(10<31>) = 0.98480
if ( 0 < n_dot_v && n_dot_v < 0.98480f)
fLift = 0.5f * kLF * medium.m_density * rel_v_len * tri_area * btSqrt(1.0f-n_dot_v*n_dot_v) * (nrm.cross(rel_v_nrm).cross(rel_v_nrm));
n.m_f += fDrag;
n.m_f += fLift;
}
else if (m_cfg.aeromodel == btSoftBody::eAeroModel::V_Point || m_cfg.aeromodel == btSoftBody::eAeroModel::V_OneSided || m_cfg.aeromodel == btSoftBody::eAeroModel::V_TwoSided)
{
if (btSoftBody::eAeroModel::V_TwoSided)
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
const btScalar dvn = btDot(rel_v,nrm);
/* Compute forces */
if(dvn>0)
{
btVector3 force(0,0,0);
const btScalar c0 = n.m_area * dvn * rel_v2/2;
const btScalar c1 = c0 * medium.m_density;
force += nrm*(-c1*kLF);
force += rel_v.normalized() * (-c1 * kDG);
ApplyClampedForce(n, force, dt);
}
}
}
}
addAeroForceToNode(m_windVelocity, i);
}
/* Pressure */
if(as_pressure)
@@ -2761,75 +2870,14 @@ void btSoftBody::applyForces()
}
}
}
/* Per face forces */
for(i=0,ni=m_faces.size();i<ni;++i)
{
btSoftBody::Face& f=m_faces[i];
if(as_faero)
{
const btVector3 v=(f.m_n[0]->m_v+f.m_n[1]->m_v+f.m_n[2]->m_v)/3;
const btVector3 x=(f.m_n[0]->m_x+f.m_n[1]->m_x+f.m_n[2]->m_x)/3;
EvaluateMedium(m_worldInfo,x,medium);
medium.m_velocity = m_windVelocity;
medium.m_density = m_worldInfo->air_density;
const btVector3 rel_v=v-medium.m_velocity;
const btScalar rel_v_len = rel_v.length();
const btScalar rel_v2=rel_v.length2();
if(rel_v2>SIMD_EPSILON)
{
const btVector3 rel_v_nrm = rel_v.normalized();
btVector3 nrm = f.m_normal;
if (m_cfg.aeromodel == btSoftBody::eAeroModel::F_TwoSidedLiftDrag)
{
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
btVector3 fDrag(0, 0, 0);
btVector3 fLift(0, 0, 0);
btScalar n_dot_v = nrm.dot(rel_v_nrm);
btScalar tri_area = 0.5f * f.m_ra;
fDrag = 0.5f * kDG * medium.m_density * rel_v2 * tri_area * n_dot_v * (-rel_v_nrm);
// Check angle of attack
// cos(10<31>) = 0.98480
if ( 0 < n_dot_v && n_dot_v < 0.98480f)
fLift = 0.5f * kLF * medium.m_density * rel_v_len * tri_area * btSqrt(1.0f-n_dot_v*n_dot_v) * (nrm.cross(rel_v_nrm).cross(rel_v_nrm));
fDrag /= 3;
fLift /= 3;
for(int j=0;j<3;++j)
{
if (f.m_n[j]->m_im>0)
{
f.m_n[j]->m_f += fDrag;
f.m_n[j]->m_f += fLift;
}
}
}
else if (m_cfg.aeromodel == btSoftBody::eAeroModel::F_OneSided || m_cfg.aeromodel == btSoftBody::eAeroModel::F_TwoSided)
{
if (btSoftBody::eAeroModel::F_TwoSided)
nrm *= (btScalar)( (btDot(nrm,rel_v) < 0) ? -1 : +1);
const btScalar dvn=btDot(rel_v,nrm);
/* Compute forces */
if(dvn>0)
{
btVector3 force(0,0,0);
const btScalar c0 = f.m_ra*dvn*rel_v2;
const btScalar c1 = c0*medium.m_density;
force += nrm*(-c1*kLF);
force += rel_v.normalized()*(-c1*kDG);
force /= 3;
for(int j=0;j<3;++j) ApplyClampedForce(*f.m_n[j],force,dt);
}
}
}
}
/* Aerodynamics */
addAeroForceToFace(m_windVelocity, i);
}
}
@@ -3276,7 +3324,7 @@ const char* btSoftBody::serialize(void* dataBuffer, class btSerializer* serializ
sbd->m_config.m_softRigidClusterImpulseSplit = m_cfg.kSR_SPLT_CL;
sbd->m_config.m_softKineticClusterImpulseSplit = m_cfg.kSK_SPLT_CL;
sbd->m_config.m_softSoftClusterImpulseSplit = m_cfg.kSS_SPLT_CL;
//pose for shape matching
{
sbd->m_pose = (SoftBodyPoseData*)serializer->getUniquePointer((void*)&m_pose);

6
src/BulletSoftBody/btSoftBody.h
View File

@@ -771,6 +771,12 @@ public:
/* Add force (or gravity) to a node of the body */
void addForce( const btVector3& force,
int node);
/* Add aero force to a node of the body */
void addAeroForceToNode(const btVector3& windVelocity,int nodeIndex);
/* Add aero force to a face of the body */
void addAeroForceToFace(const btVector3& windVelocity,int faceIndex);
/* Add velocity to the entire body */
void addVelocity( const btVector3& velocity);

24
src/LinearMath/btConvexHullComputer.cpp
View File

@@ -272,8 +272,8 @@ class btConvexHullInternal
class Rational64
{
private:
uint64_t numerator;
uint64_t denominator;
uint64_t m_numerator;
uint64_t m_denominator;
int sign;
public:
@@ -282,48 +282,48 @@ class btConvexHullInternal
if (numerator > 0)
{
sign = 1;
this->numerator = (uint64_t) numerator;
m_numerator = (uint64_t) numerator;
}
else if (numerator < 0)
{
sign = -1;
this->numerator = (uint64_t) -numerator;
m_numerator = (uint64_t) -numerator;
}
else
{
sign = 0;
this->numerator = 0;
m_numerator = 0;
}
if (denominator > 0)
{
this->denominator = (uint64_t) denominator;
m_denominator = (uint64_t) denominator;
}
else if (denominator < 0)
{
sign = -sign;
this->denominator = (uint64_t) -denominator;
m_denominator = (uint64_t) -denominator;
}
else
{
this->denominator = 0;
m_denominator = 0;
}
}
bool isNegativeInfinity() const
{
return (sign < 0) && (denominator == 0);
return (sign < 0) && (m_denominator == 0);
}
bool isNaN() const
{
return (sign == 0) && (denominator == 0);
return (sign == 0) && (m_denominator == 0);
}
int compare(const Rational64& b) const;
btScalar toScalar() const
{
return sign * ((denominator == 0) ? SIMD_INFINITY : (btScalar) numerator / denominator);
return sign * ((m_denominator == 0) ? SIMD_INFINITY : (btScalar) m_numerator / m_denominator);
}
};
@@ -932,7 +932,7 @@ int btConvexHullInternal::Rational64::compare(const Rational64& b) const
#else
return sign * Int128::mul(numerator, b.denominator).ucmp(Int128::mul(denominator, b.numerator));
return sign * Int128::mul(m_numerator, b.m_denominator).ucmp(Int128::mul(m_denominator, b.m_numerator));
#endif
}

14
src/MiniCL/MiniCL.cpp
View File

@@ -73,6 +73,20 @@ CL_API_ENTRY cl_int CL_API_CALL clGetPlatformInfo(
}
switch(param_name)
{
case CL_PLATFORM_VERSION:
{
if(param_value_size < (strlen(spDriverVersion) + 1))
{
return CL_INVALID_VALUE;
}
strcpy((char*)param_value, spDriverVersion);
if(param_value_size_ret != NULL)
{
*param_value_size_ret = strlen(spDriverVersion) + 1;
}
break;
}
case CL_PLATFORM_NAME:
case CL_PLATFORM_VENDOR :
if(param_value_size < (strlen(spPlatformID) + 1))
{