From 0d1fcf7c48a01aa59b659feb22ae9d71197c5770 Mon Sep 17 00:00:00 2001 From: "erwin.coumans" Date: Mon, 14 Nov 2011 22:22:49 +0000 Subject: [PATCH] Move btOclUtils.cpp/h and btOclCommon.cpp/h to btOpenCLUtils.cpp/h 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 --- Demos/OpenCLClothDemo/AMD/CMakeLists.txt | 7 +- Demos/OpenCLClothDemo/AMD/premake4.lua | 7 +- Demos/OpenCLClothDemo/Apple/CMakeLists.txt | 7 +- Demos/OpenCLClothDemo/Intel/CMakeLists.txt | 7 +- Demos/OpenCLClothDemo/Intel/premake4.lua | 7 +- Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt | 7 +- Demos/OpenCLClothDemo/NVidia/CMakeLists.txt | 7 +- Demos/OpenCLClothDemo/NVidia/premake4.lua | 7 +- Demos/OpenCLClothDemo/clstuff.cpp | 45 +- Demos/ParticlesOpenCL/AMD/CMakeLists.txt | 7 +- Demos/ParticlesOpenCL/Apple/CMakeLists.txt | 7 +- Demos/ParticlesOpenCL/Intel/CMakeLists.txt | 7 +- Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt | 7 +- Demos/ParticlesOpenCL/NVidia/CMakeLists.txt | 7 +- .../btParticlesDemoDynamicsWorld.cpp | 32 +- Demos/SerializeDemo/AMD/CMakeLists.txt | 24 +- Demos/SerializeDemo/main.cpp | 37 +- Demos/SharedOpenCL/btOclCommon.cpp | 121 ----- Demos/SharedOpenCL/btOclUtils.cpp | 284 ----------- Demos/SharedOpenCL/btOclUtils.h | 42 -- .../{btOclCommon.h => btOpenCLInclude.h} | 44 +- Demos/SharedOpenCL/btOpenCLUtils.cpp | 451 ++++++++++++++++++ Demos/SharedOpenCL/btOpenCLUtils.h | 101 ++++ src/BulletSoftBody/btSoftBody.cpp | 284 ++++++----- src/BulletSoftBody/btSoftBody.h | 6 + src/LinearMath/btConvexHullComputer.cpp | 24 +- src/MiniCL/MiniCL.cpp | 14 + 27 files changed, 877 insertions(+), 723 deletions(-) delete mode 100644 Demos/SharedOpenCL/btOclCommon.cpp delete mode 100644 Demos/SharedOpenCL/btOclUtils.cpp delete mode 100644 Demos/SharedOpenCL/btOclUtils.h rename Demos/SharedOpenCL/{btOclCommon.h => btOpenCLInclude.h} (56%) create mode 100644 Demos/SharedOpenCL/btOpenCLUtils.cpp create mode 100644 Demos/SharedOpenCL/btOpenCLUtils.h diff --git a/Demos/OpenCLClothDemo/AMD/CMakeLists.txt b/Demos/OpenCLClothDemo/AMD/CMakeLists.txt index 391166b4f..1795f83c0 100644 --- a/Demos/OpenCLClothDemo/AMD/CMakeLists.txt +++ b/Demos/OpenCLClothDemo/AMD/CMakeLists.txt @@ -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 diff --git a/Demos/OpenCLClothDemo/AMD/premake4.lua b/Demos/OpenCLClothDemo/AMD/premake4.lua index 2f59b49e5..edaf53acc 100644 --- a/Demos/OpenCLClothDemo/AMD/premake4.lua +++ b/Demos/OpenCLClothDemo/AMD/premake4.lua @@ -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", diff --git a/Demos/OpenCLClothDemo/Apple/CMakeLists.txt b/Demos/OpenCLClothDemo/Apple/CMakeLists.txt index 629a8dffc..40363be82 100644 --- a/Demos/OpenCLClothDemo/Apple/CMakeLists.txt +++ b/Demos/OpenCLClothDemo/Apple/CMakeLists.txt @@ -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 diff --git a/Demos/OpenCLClothDemo/Intel/CMakeLists.txt b/Demos/OpenCLClothDemo/Intel/CMakeLists.txt index 0018c09a5..0acc2c5be 100644 --- a/Demos/OpenCLClothDemo/Intel/CMakeLists.txt +++ b/Demos/OpenCLClothDemo/Intel/CMakeLists.txt @@ -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 diff --git a/Demos/OpenCLClothDemo/Intel/premake4.lua b/Demos/OpenCLClothDemo/Intel/premake4.lua index 9adc7ffdd..d2f9c1088 100644 --- a/Demos/OpenCLClothDemo/Intel/premake4.lua +++ b/Demos/OpenCLClothDemo/Intel/premake4.lua @@ -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", diff --git a/Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt b/Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt index a4d3b5b7b..9470e1c3d 100644 --- a/Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt +++ b/Demos/OpenCLClothDemo/MiniCL/CMakeLists.txt @@ -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) diff --git a/Demos/OpenCLClothDemo/NVidia/CMakeLists.txt b/Demos/OpenCLClothDemo/NVidia/CMakeLists.txt index 2c6f12840..565bb047c 100644 --- a/Demos/OpenCLClothDemo/NVidia/CMakeLists.txt +++ b/Demos/OpenCLClothDemo/NVidia/CMakeLists.txt @@ -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 diff --git a/Demos/OpenCLClothDemo/NVidia/premake4.lua b/Demos/OpenCLClothDemo/NVidia/premake4.lua index 8747f30b9..4a0f36c35 100644 --- a/Demos/OpenCLClothDemo/NVidia/premake4.lua +++ b/Demos/OpenCLClothDemo/NVidia/premake4.lua @@ -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", diff --git a/Demos/OpenCLClothDemo/clstuff.cpp b/Demos/OpenCLClothDemo/clstuff.cpp index 13692139e..47cc7d65b 100644 --- a/Demos/OpenCLClothDemo/clstuff.cpp +++ b/Demos/OpenCLClothDemo/clstuff.cpp @@ -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 @@ -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); diff --git a/Demos/ParticlesOpenCL/AMD/CMakeLists.txt b/Demos/ParticlesOpenCL/AMD/CMakeLists.txt index 510a63132..6dfe27b88 100644 --- a/Demos/ParticlesOpenCL/AMD/CMakeLists.txt +++ b/Demos/ParticlesOpenCL/AMD/CMakeLists.txt @@ -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 diff --git a/Demos/ParticlesOpenCL/Apple/CMakeLists.txt b/Demos/ParticlesOpenCL/Apple/CMakeLists.txt index de9605343..22ecedbd8 100644 --- a/Demos/ParticlesOpenCL/Apple/CMakeLists.txt +++ b/Demos/ParticlesOpenCL/Apple/CMakeLists.txt @@ -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) diff --git a/Demos/ParticlesOpenCL/Intel/CMakeLists.txt b/Demos/ParticlesOpenCL/Intel/CMakeLists.txt index 27e9f3b70..54828a5cc 100644 --- a/Demos/ParticlesOpenCL/Intel/CMakeLists.txt +++ b/Demos/ParticlesOpenCL/Intel/CMakeLists.txt @@ -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 diff --git a/Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt b/Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt index 2d5eef1aa..7e57ef3a7 100644 --- a/Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt +++ b/Demos/ParticlesOpenCL/MiniCL/CMakeLists.txt @@ -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 ) diff --git a/Demos/ParticlesOpenCL/NVidia/CMakeLists.txt b/Demos/ParticlesOpenCL/NVidia/CMakeLists.txt index 727256aba..8013a970f 100644 --- a/Demos/ParticlesOpenCL/NVidia/CMakeLists.txt +++ b/Demos/ParticlesOpenCL/NVidia/CMakeLists.txt @@ -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 diff --git a/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp b/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp index ab3183fc9..24b4636f2 100644 --- a/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp +++ b/Demos/ParticlesOpenCL/btParticlesDemoDynamicsWorld.cpp @@ -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); diff --git a/Demos/SerializeDemo/AMD/CMakeLists.txt b/Demos/SerializeDemo/AMD/CMakeLists.txt index 38ae8501f..abe46ebac 100644 --- a/Demos/SerializeDemo/AMD/CMakeLists.txt +++ b/Demos/SerializeDemo/AMD/CMakeLists.txt @@ -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) diff --git a/Demos/SerializeDemo/main.cpp b/Demos/SerializeDemo/main.cpp index 5665975ce..cd5914718 100644 --- a/Demos/SerializeDemo/main.cpp +++ b/Demos/SerializeDemo/main.cpp @@ -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 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 diff --git a/Demos/SharedOpenCL/btOclCommon.cpp b/Demos/SharedOpenCL/btOclCommon.cpp deleted file mode 100644 index ed8180f6d..000000000 --- a/Demos/SharedOpenCL/btOclCommon.cpp +++ /dev/null @@ -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 - -#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; -} - - diff --git a/Demos/SharedOpenCL/btOclUtils.cpp b/Demos/SharedOpenCL/btOclUtils.cpp deleted file mode 100644 index 80918e982..000000000 --- a/Demos/SharedOpenCL/btOclUtils.cpp +++ /dev/null @@ -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 -#include -#include - - -#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 \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_ - myprintf(" CL_DEVICE_PREFERRED_VECTOR_WIDTH_\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]); -} - diff --git a/Demos/SharedOpenCL/btOclUtils.h b/Demos/SharedOpenCL/btOclUtils.h deleted file mode 100644 index 1636dd2b1..000000000 --- a/Demos/SharedOpenCL/btOclUtils.h +++ /dev/null @@ -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 -#else //USE_MINICL -#ifdef BT_USE_CLEW - #include "clew.h" -#else -#ifdef __APPLE__ - #include -#else - #include -#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 diff --git a/Demos/SharedOpenCL/btOclCommon.h b/Demos/SharedOpenCL/btOpenCLInclude.h similarity index 56% rename from Demos/SharedOpenCL/btOclCommon.h rename to Demos/SharedOpenCL/btOpenCLInclude.h index 7248adb24..830d6277b 100644 --- a/Demos/SharedOpenCL/btOclCommon.h +++ b/Demos/SharedOpenCL/btOpenCLInclude.h @@ -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 +#else +#include +#endif +#else +#ifdef USE_MINICL +#include +#else +#include +#ifdef _WIN32 +#include "CL/cl_gl.h" +#endif //_WIN32 +#endif +#endif //__APPLE__ + +#include +#include +#define oclCHECKERROR(a, b) if((a)!=(b)) { printf("OCL Error : %d\n", (a)); assert((a) == (b)); } + + +#endif //BT_OPENCL_INCLUDE_H -#endif // BTOCLCOMMON_H diff --git a/Demos/SharedOpenCL/btOpenCLUtils.cpp b/Demos/SharedOpenCL/btOpenCLUtils.cpp new file mode 100644 index 000000000..0e79acdc2 --- /dev/null +++ b/Demos/SharedOpenCL/btOpenCLUtils.cpp @@ -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 + +#include "btOpenCLUtils.h" +#include +#include + + +//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 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 "); + 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"); + 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_ + 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; + +} diff --git a/Demos/SharedOpenCL/btOpenCLUtils.h b/Demos/SharedOpenCL/btOpenCLUtils.h new file mode 100644 index 000000000..52b749762 --- /dev/null +++ b/Demos/SharedOpenCL/btOpenCLUtils.h @@ -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 diff --git a/src/BulletSoftBody/btSoftBody.cpp b/src/BulletSoftBody/btSoftBody.cpp index 0d7d98b3a..d1b5eb43d 100644 --- a/src/BulletSoftBody/btSoftBody.cpp +++ b/src/BulletSoftBody/btSoftBody.cpp @@ -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º) = 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º) = 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º) = 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();im_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º) = 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); diff --git a/src/BulletSoftBody/btSoftBody.h b/src/BulletSoftBody/btSoftBody.h index 79bde1dfd..ba589486f 100644 --- a/src/BulletSoftBody/btSoftBody.h +++ b/src/BulletSoftBody/btSoftBody.h @@ -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); diff --git a/src/LinearMath/btConvexHullComputer.cpp b/src/LinearMath/btConvexHullComputer.cpp index 2d7c22975..b47cb81f2 100644 --- a/src/LinearMath/btConvexHullComputer.cpp +++ b/src/LinearMath/btConvexHullComputer.cpp @@ -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 } diff --git a/src/MiniCL/MiniCL.cpp b/src/MiniCL/MiniCL.cpp index 4307f63ca..e998f1dfe 100644 --- a/src/MiniCL/MiniCL.cpp +++ b/src/MiniCL/MiniCL.cpp @@ -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)) {