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

Merge branch 'master' of https://bitbucket.org/erwincoumans/bullet3_experiments

Browse Source
This commit is contained in:
Erwin Coumans
2013-04-29 22:55:38 -07:00
parent 13e134aba9 a5bba9f792
commit 6b64517817
248 changed files with 6798 additions and 9555 deletions
Download Patch File Download Diff File Expand all files Collapse all files

8
demo/gpudemo/GpuDemo.cpp → Demos3/GpuDemos/GpuDemo.cpp
View File

@@ -1,7 +1,7 @@
#include "GpuDemo.h"
#include "GpuDemoInternalData.h"
#include "Bullet3Common/b3Scalar.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/ShapeData.h"
#include "OpenGLWindow/GLInstancingRenderer.h"
@@ -15,7 +15,7 @@ GpuDemo::~GpuDemo()
{
if (m_clData)
{
btAssert(m_clData->m_clInitialized==false);
b3Assert(m_clData->m_clInitialized==false);
delete m_clData;
}
@@ -69,7 +69,7 @@ void GpuDemo::initCL(int preferredDeviceIndex, int preferredPlatformIndex)
oclCHECKERROR(ciErrNum, CL_SUCCESS);
b3OpenCLUtils::printDeviceInfo(m_clData->m_clDevice);
btOpenCLDeviceInfo info;
b3OpenCLDeviceInfo info;
b3OpenCLUtils::getDeviceInfo(m_clData->m_clDevice,&info);
m_clData->m_clDeviceName = info.m_deviceName;
m_clData->m_clInitialized = true;
@@ -98,7 +98,7 @@ int GpuDemo::registerGraphicsSphereShape(const ConstructionInfo& ci, float radiu
{
int numVertices = sizeof(point_sphere_vertices)/strideInBytes;
int numIndices = sizeof(point_sphere_indices)/sizeof(int);
graphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,BT_GL_POINTS);
graphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,B3_GL_POINTS);
} else
{
if (radius>=mediumSphereThreshold)

2
demo/gpudemo/GpuDemo.h → Demos3/GpuDemos/GpuDemo.h
View File

@@ -31,7 +31,7 @@ public:
float gapY;
float gapZ;
GLInstancingRenderer* m_instancingRenderer;
class btgWindowInterface* m_window;
class b3gWindowInterface* m_window;
class GwenUserInterface* m_gui;
ConstructionInfo()

2
demo/gpudemo/GpuDemoInternalData.h → Demos3/GpuDemos/GpuDemoInternalData.h
View File

@@ -1,7 +1,7 @@
#ifndef GPU_DEMO_INTERNAL_DATA_H
#define GPU_DEMO_INTERNAL_DATA_H
#include "basic_initialize/b3OpenCLInclude.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLInclude.h"
struct GpuDemoInternalData
{

74
demo/gpudemo/ParticleDemo.cpp → Demos3/GpuDemos/ParticleDemo.cpp
View File

@@ -2,7 +2,7 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "OpenGLWindow/ShapeData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#define MSTRINGIFY(A) #A
static char* particleKernelsString =
@@ -12,10 +12,10 @@ static char* particleKernelsString =
#include "Bullet3Common/b3Vector3.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
//#include "../../opencl/primitives/AdlPrimitives/Math/Math.h"
//#include "../../opencl/broadphase_benchmark/btGridBroadphaseCL.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
//#include "../../opencl/broadphase_benchmark/b3GridBroadphaseCL.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "GpuDemoInternalData.h"
@@ -43,9 +43,9 @@ static char* particleKernelsString =
ATTRIBUTE_ALIGNED16(struct) btSimParams
B3_ATTRIBUTE_ALIGNED16(struct) b3SimParams
{
BT_DECLARE_ALIGNED_ALLOCATOR();
B3_DECLARE_ALIGNED_ALLOCATOR();
b3Vector3 m_gravity;
float m_worldMin[4];
float m_worldMax[4];
@@ -61,7 +61,7 @@ ATTRIBUTE_ALIGNED16(struct) btSimParams
float m_dummy;
btSimParams()
b3SimParams()
{
m_gravity.setValue(0,-0.03,0.f);
m_particleRad = 0.023f;
@@ -98,10 +98,10 @@ struct ParticleInternalData
cl_mem m_clPositionBuffer;
b3AlignedObjectArray<b3Vector3> m_velocitiesCPU;
btOpenCLArray<b3Vector3>* m_velocitiesGPU;
b3OpenCLArray<b3Vector3>* m_velocitiesGPU;
b3AlignedObjectArray<btSimParams> m_simParamCPU;
btOpenCLArray<btSimParams>* m_simParamGPU;
b3AlignedObjectArray<b3SimParams> m_simParamCPU;
b3OpenCLArray<b3SimParams>* m_simParamGPU;
@@ -170,12 +170,12 @@ void ParticleDemo::setupScene(const ConstructionInfo& ci)
m_data->m_broadphaseGPU = new b3GpuSapBroadphase(m_clData->m_clContext ,m_clData->m_clDevice,m_clData->m_clQueue);//overlappingPairCache,b3Vector3(4.f, 4.f, 4.f), 128, 128, 128,maxObjects, maxObjects, maxPairsSmallProxy, 100.f, 128,
/*m_data->m_broadphaseGPU = new btGridBroadphaseCl(overlappingPairCache,b3Vector3(radius,radius,radius), 128, 128, 128,
/*m_data->m_broadphaseGPU = new b3GridBroadphaseCl(overlappingPairCache,b3Vector3(radius,radius,radius), 128, 128, 128,
maxObjects, maxObjects, maxPairsSmallProxy, 100.f, 128,
m_clData->m_clContext ,m_clData->m_clDevice,m_clData->m_clQueue);
*/
m_data->m_velocitiesGPU = new btOpenCLArray<b3Vector3>(m_clData->m_clContext,m_clData->m_clQueue,numParticles);
m_data->m_velocitiesGPU = new b3OpenCLArray<b3Vector3>(m_clData->m_clContext,m_clData->m_clQueue,numParticles);
m_data->m_velocitiesCPU.resize(numParticles);
for (int i=0;i<numParticles;i++)
{
@@ -183,7 +183,7 @@ void ParticleDemo::setupScene(const ConstructionInfo& ci)
}
m_data->m_velocitiesGPU->copyFromHost(m_data->m_velocitiesCPU);
m_data->m_simParamGPU = new btOpenCLArray<btSimParams>(m_clData->m_clContext,m_clData->m_clQueue,1,false);
m_data->m_simParamGPU = new b3OpenCLArray<b3SimParams>(m_clData->m_clContext,m_clData->m_clQueue,1,false);
m_data->m_simParamGPU->copyFromHost(m_data->m_simParamCPU);
cl_int pErrNum;
@@ -210,7 +210,7 @@ void ParticleDemo::setupScene(const ConstructionInfo& ci)
{
int numVertices = sizeof(point_sphere_vertices)/strideInBytes;
int numIndices = sizeof(point_sphere_indices)/sizeof(int);
shapeId = m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,BT_GL_POINTS);
shapeId = m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,B3_GL_POINTS);
} else
{
int numVertices = sizeof(low_sphere_vertices)/strideInBytes;
@@ -341,14 +341,14 @@ void ParticleDemo::clientMoveAndDisplay()
if (0)
{
btBufferInfoCL bInfo[] = {
btBufferInfoCL( m_data->m_velocitiesGPU->getBufferCL(), true ),
btBufferInfoCL( m_data->m_clPositionBuffer)
b3BufferInfoCL bInfo[] = {
b3BufferInfoCL( m_data->m_velocitiesGPU->getBufferCL(), true ),
b3BufferInfoCL( m_data->m_clPositionBuffer)
};
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_updatePositionsKernel );
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_updatePositionsKernel );
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(btBufferInfoCL) );
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(b3BufferInfoCL) );
launcher.setConst( numParticles);
launcher.launch1D( numParticles);
@@ -359,16 +359,16 @@ void ParticleDemo::clientMoveAndDisplay()
if (1)
{
btBufferInfoCL bInfo[] = {
btBufferInfoCL( m_data->m_clPositionBuffer),
btBufferInfoCL( m_data->m_velocitiesGPU->getBufferCL() ),
btBufferInfoCL( m_data->m_simParamGPU->getBufferCL(),true)
b3BufferInfoCL bInfo[] = {
b3BufferInfoCL( m_data->m_clPositionBuffer),
b3BufferInfoCL( m_data->m_velocitiesGPU->getBufferCL() ),
b3BufferInfoCL( m_data->m_simParamGPU->getBufferCL(),true)
};
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_updatePositionsKernel2 );
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_updatePositionsKernel2 );
launcher.setConst( numParticles);
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(btBufferInfoCL) );
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(b3BufferInfoCL) );
float timeStep = 1.f/60.f;
launcher.setConst( timeStep);
@@ -378,13 +378,13 @@ void ParticleDemo::clientMoveAndDisplay()
}
{
btBufferInfoCL bInfo[] = {
btBufferInfoCL( m_data->m_clPositionBuffer),
btBufferInfoCL( m_data->m_broadphaseGPU->getAabbBufferWS()),
b3BufferInfoCL bInfo[] = {
b3BufferInfoCL( m_data->m_clPositionBuffer),
b3BufferInfoCL( m_data->m_broadphaseGPU->getAabbBufferWS()),
};
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_updateAabbsKernel );
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(btBufferInfoCL) );
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_updateAabbsKernel );
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(b3BufferInfoCL) );
launcher.setConst( m_data->m_simParamCPU[0].m_particleRad);
launcher.setConst( numParticles);
@@ -404,14 +404,14 @@ void ParticleDemo::clientMoveAndDisplay()
if (numPairsGPU)
{
btBufferInfoCL bInfo[] = {
btBufferInfoCL( m_data->m_clPositionBuffer),
btBufferInfoCL( m_data->m_velocitiesGPU->getBufferCL() ),
btBufferInfoCL( m_data->m_broadphaseGPU->getOverlappingPairBuffer(),true),
b3BufferInfoCL bInfo[] = {
b3BufferInfoCL( m_data->m_clPositionBuffer),
b3BufferInfoCL( m_data->m_velocitiesGPU->getBufferCL() ),
b3BufferInfoCL( m_data->m_broadphaseGPU->getOverlappingPairBuffer(),true),
};
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_collideParticlesKernel);
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(btBufferInfoCL) );
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_collideParticlesKernel);
launcher.setBuffers( bInfo, sizeof(bInfo)/sizeof(b3BufferInfoCL) );
launcher.setConst( numPairsGPU);
launcher.launch1D( numPairsGPU);
clFinish(m_clData->m_clQueue);
@@ -454,7 +454,7 @@ void ParticleDemo::clientMoveAndDisplay()
glFlush();
char* orgBase = (char*)glMapBuffer( GL_ARRAY_BUFFER,GL_READ_WRITE);
//btGraphicsInstance* gfxObj = m_graphicsInstances[k];
//b3GraphicsInstance* gfxObj = m_graphicsInstances[k];
int totalNumInstances= numParticles;

0
demo/gpudemo/ParticleDemo.h → Demos3/GpuDemos/ParticleDemo.h
View File

0
demo/gpudemo/ParticleKernels.cl → Demos3/GpuDemos/ParticleKernels.cl
View File

46
demo/gpudemo/broadphase/PairBench.cpp → Demos3/GpuDemos/broadphase/PairBench.cpp
View File

@@ -3,15 +3,15 @@
#include "OpenGLWindow/ShapeData.h"
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
static btKeyboardCallback oldCallback = 0;
static b3KeyboardCallback oldCallback = 0;
extern bool gReset;
#define MSTRINGIFY(A) #A
@@ -70,9 +70,9 @@ typedef struct
float fy;
float fz;
int uw;
} btAABBCL;
} b3AABBCL;
__kernel void updateAabbSimple( __global float4* posOrnColors, const int numNodes, __global btAABBCL* pAABB)
__kernel void updateAabbSimple( __global float4* posOrnColors, const int numNodes, __global b3AABBCL* pAABB)
{
int nodeId = get_global_id(0);
if( nodeId < numNodes )
@@ -102,8 +102,8 @@ struct PairBenchInternalData
cl_kernel m_colorPairsKernel;
cl_kernel m_updateAabbSimple;
btOpenCLArray<btVector4>* m_instancePosOrnColor;
btOpenCLArray<float>* m_bodyTimes;
b3OpenCLArray<b3Vector4>* m_instancePosOrnColor;
b3OpenCLArray<float>* m_bodyTimes;
PairBenchInternalData()
:m_broadphaseGPU(0),
m_moveObjectsKernel(0),
@@ -141,7 +141,7 @@ static void PairKeyboardCallback(int key, int state)
gReset = true;
}
//btDefaultKeyboardCallback(key,state);
//b3DefaultKeyboardCallback(key,state);
oldCallback(key,state);
}
@@ -172,7 +172,7 @@ void PairBench::initPhysics(const ConstructionInfo& ci)
m_instancingRenderer = ci.m_instancingRenderer;
CProfileManager::CleanupMemory();
b3ProfileManager::CleanupMemory();
int strideInBytes = 9*sizeof(float);
int numVertices = sizeof(cube_vertices)/strideInBytes;
int numIndices = sizeof(cube_vertices)/sizeof(int);
@@ -190,8 +190,8 @@ void PairBench::initPhysics(const ConstructionInfo& ci)
b3Vector3 position(k*3,i*3,j*3);
b3Quaternion orn(0,0,0,1);
btVector4 color(0,1,0,1);
btVector4 scaling(1,1,1,1);
b3Vector4 color(0,1,0,1);
b3Vector4 scaling(1,1,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
b3Vector3 aabbHalfExtents(1,1,1);
@@ -236,21 +236,21 @@ void PairBench::clientMoveAndDisplay()
bool animate=true;
int numObjects= m_instancingRenderer->getInternalData()->m_totalNumInstances;
btVector4* positions = 0;
b3Vector4* positions = 0;
if (animate)
{
GLuint vbo = m_instancingRenderer->getInternalData()->m_vbo;
int arraySizeInBytes = numObjects * (3)*sizeof(btVector4);
int arraySizeInBytes = numObjects * (3)*sizeof(b3Vector4);
glBindBuffer(GL_ARRAY_BUFFER, vbo);
cl_bool blocking= CL_TRUE;
char* hostPtr= (char*)glMapBufferRange( GL_ARRAY_BUFFER,m_instancingRenderer->getMaxShapeCapacity(),arraySizeInBytes, GL_MAP_WRITE_BIT|GL_MAP_READ_BIT );//GL_READ_WRITE);//GL_WRITE_ONLY
GLint err = glGetError();
assert(err==GL_NO_ERROR);
positions = (btVector4*)hostPtr;
positions = (b3Vector4*)hostPtr;
if (m_data->m_instancePosOrnColor && m_data->m_instancePosOrnColor->size() != 3*numObjects)
{
@@ -259,10 +259,10 @@ void PairBench::clientMoveAndDisplay()
}
if (!m_data->m_instancePosOrnColor)
{
m_data->m_instancePosOrnColor = new btOpenCLArray<btVector4>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_instancePosOrnColor = new b3OpenCLArray<b3Vector4>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_instancePosOrnColor->resize(3*numObjects);
m_data->m_instancePosOrnColor->copyFromHostPointer(positions,3*numObjects,0);
m_data->m_bodyTimes = new btOpenCLArray<float>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_bodyTimes = new b3OpenCLArray<float>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_bodyTimes ->resize(numObjects);
b3AlignedObjectArray<float> tmp;
tmp.resize(numObjects);
@@ -278,7 +278,7 @@ void PairBench::clientMoveAndDisplay()
if (1)
{
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_sineWaveKernel);
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_sineWaveKernel);
launcher.setBuffer(m_data->m_instancePosOrnColor->getBufferCL() );
launcher.setBuffer(m_data->m_bodyTimes->getBufferCL() );
launcher.setConst( numObjects);
@@ -288,7 +288,7 @@ void PairBench::clientMoveAndDisplay()
else
{
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_moveObjectsKernel);
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_moveObjectsKernel);
launcher.setBuffer(m_data->m_instancePosOrnColor->getBufferCL() );
launcher.setConst( numObjects);
launcher.launch1D( numObjects);
@@ -298,7 +298,7 @@ void PairBench::clientMoveAndDisplay()
}
{
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_updateAabbSimple);
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_updateAabbSimple);
launcher.setBuffer(m_data->m_instancePosOrnColor->getBufferCL() );
launcher.setConst( numObjects);
launcher.setBuffer(m_data->m_broadphaseGPU->getAabbBufferWS());
@@ -307,7 +307,7 @@ void PairBench::clientMoveAndDisplay()
}
{
BT_PROFILE("calculateOverlappingPairs");
B3_PROFILE("calculateOverlappingPairs");
m_data->m_broadphaseGPU->calculateOverlappingPairs();
//int numPairs = m_data->m_broadphaseGPU->getNumOverlap();
//printf("numPairs = %d\n", numPairs);
@@ -331,7 +331,7 @@ void PairBench::clientMoveAndDisplay()
int numPairs = m_data->m_broadphaseGPU->getNumOverlap();
cl_mem pairBuf = m_data->m_broadphaseGPU->getOverlappingPairBuffer();
btLauncherCL launcher(m_clData->m_clQueue, m_data->m_colorPairsKernel);
b3LauncherCL launcher(m_clData->m_clQueue, m_data->m_colorPairsKernel);
launcher.setBuffer(m_data->m_instancePosOrnColor->getBufferCL() );
launcher.setConst( numObjects);
launcher.setBuffer( pairBuf);

2
demo/gpudemo/broadphase/PairBench.h → Demos3/GpuDemos/broadphase/PairBench.h
View File

@@ -7,7 +7,7 @@ class PairBench : public GpuDemo
{
class GLInstancingRenderer* m_instancingRenderer;
class btgWindowInterface* m_window;
class b3gWindowInterface* m_window;
struct PairBenchInternalData* m_data;

8
demo/gpudemo/gwenUserInterface.cpp → Demos3/GpuDemos/gwenUserInterface.cpp
View File

@@ -39,8 +39,8 @@ struct GwenInternalData
Gwen::Controls::Label* m_leftStatusBar;
b3AlignedObjectArray<struct Gwen::Event::Handler*> m_handlers;
btToggleButtonCallback m_toggleButtonCallback;
btComboBoxCallback m_comboBoxCallback;
b3ToggleButtonCallback m_toggleButtonCallback;
b3ComboBoxCallback m_comboBoxCallback;
};
GwenUserInterface::GwenUserInterface()
@@ -243,7 +243,7 @@ void GwenUserInterface::init(int width, int height,struct sth_stash* stash,float
}
void GwenUserInterface::setToggleButtonCallback(btToggleButtonCallback callback)
void GwenUserInterface::setToggleButtonCallback(b3ToggleButtonCallback callback)
{
m_data->m_toggleButtonCallback = callback;
}
@@ -269,7 +269,7 @@ void GwenUserInterface::registerToggleButton(int buttonId, const char* name)
}
void GwenUserInterface::setComboBoxCallback(btComboBoxCallback callback)
void GwenUserInterface::setComboBoxCallback(b3ComboBoxCallback callback)
{
m_data->m_comboBoxCallback = callback;
}

8
demo/gpudemo/gwenUserInterface.h → Demos3/GpuDemos/gwenUserInterface.h
View File

@@ -3,8 +3,8 @@
struct GwenInternalData;
typedef void (*btComboBoxCallback) (int combobox, const char* item);
typedef void (*btToggleButtonCallback)(int button, int state);
typedef void (*b3ComboBoxCallback) (int combobox, const char* item);
typedef void (*b3ToggleButtonCallback)(int button, int state);
class GwenUserInterface
{
@@ -25,10 +25,10 @@ class GwenUserInterface
bool mouseMoveCallback( float x, float y);
bool mouseButtonCallback(int button, int state, float x, float y);
void setToggleButtonCallback(btToggleButtonCallback callback);
void setToggleButtonCallback(b3ToggleButtonCallback callback);
void registerToggleButton(int buttonId, const char* name);
void setComboBoxCallback(btComboBoxCallback callback);
void setComboBoxCallback(b3ComboBoxCallback callback);
void registerComboBox(int buttonId, int numItems, const char** items);
void setStatusBarMessage(const char* message, bool isLeft=true);

62
demo/gpudemo/main_opengl3core.cpp → Demos3/GpuDemos/main_opengl3core.cpp
View File

@@ -17,7 +17,7 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
//#include "OpenGL3CoreRenderer.h"
#include "Bullet3Common/b3Quickprof.h"
//#include "btGpuDynamicsWorld.h"
//#include "b3GpuDynamicsWorld.h"
#include <assert.h>
#include <string.h>
#include "OpenGLTrueTypeFont/fontstash.h"
@@ -44,7 +44,7 @@ static void MyResizeCallback( float width, float height)
g_OpenGLHeight = height;
}
btgWindowInterface* window=0;
b3gWindowInterface* window=0;
GwenUserInterface* gui = 0;
bool gPause = false;
bool gReset = false;
@@ -161,7 +161,7 @@ static void MyMouseMoveCallback( float x, float y)
{
bool handled = gui ->mouseMoveCallback(x,y);
if (!handled)
btDefaultMouseMoveCallback(x,y);
b3DefaultMouseMoveCallback(x,y);
}
}
static void MyMouseButtonCallback(int button, int state, float x, float y)
@@ -170,18 +170,18 @@ static void MyMouseButtonCallback(int button, int state, float x, float y)
{
bool handled = gui->mouseButtonCallback(button,state,x,y);
if (!handled)
btDefaultMouseButtonCallback(button,state,x,y);
b3DefaultMouseButtonCallback(button,state,x,y);
}
}
void MyKeyboardCallback(int key, int state)
{
if (key==BTG_ESCAPE && window)
if (key==B3G_ESCAPE && window)
{
window->setRequestExit();
}
btDefaultKeyboardCallback(key,state);
b3DefaultKeyboardCallback(key,state);
}
@@ -325,15 +325,15 @@ void Usage()
void DumpSimulationTime(FILE* f)
{
CProfileIterator* profileIterator = CProfileManager::Get_Iterator();
b3ProfileIterator* profileIterator = b3ProfileManager::Get_Iterator();
profileIterator->First();
if (profileIterator->Is_Done())
return;
float accumulated_time=0,parent_time = profileIterator->Is_Root() ? CProfileManager::Get_Time_Since_Reset() : profileIterator->Get_Current_Parent_Total_Time();
float accumulated_time=0,parent_time = profileIterator->Is_Root() ? b3ProfileManager::Get_Time_Since_Reset() : profileIterator->Get_Current_Parent_Total_Time();
int i;
int frames_since_reset = CProfileManager::Get_Frame_Count_Since_Reset();
int frames_since_reset = b3ProfileManager::Get_Frame_Count_Since_Reset();
//fprintf(f,"%.3f,", parent_time );
float totalTime = 0.f;
@@ -351,7 +351,7 @@ void DumpSimulationTime(FILE* f)
{
float current_total_time = profileIterator->Get_Current_Total_Time();
accumulated_time += current_total_time;
float fraction = parent_time > SIMD_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
float fraction = parent_time > B3_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
const char* name = profileIterator->Get_Current_Name();
fprintf(f,"%s,",name);
}
@@ -365,7 +365,7 @@ void DumpSimulationTime(FILE* f)
{
float current_total_time = profileIterator->Get_Current_Total_Time();
accumulated_time += current_total_time;
float fraction = parent_time > SIMD_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
float fraction = parent_time > B3_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
const char* name = profileIterator->Get_Current_Name();
//if (!strcmp(name,"stepSimulation"))
{
@@ -378,7 +378,7 @@ void DumpSimulationTime(FILE* f)
fprintf(f,"\n");
CProfileManager::Release_Iterator(profileIterator);
b3ProfileManager::Release_Iterator(profileIterator);
}
@@ -440,14 +440,14 @@ int main(int argc, char* argv[])
printf("Preferred cl_platform index%d\n", ci.preferredOpenCLPlatformIndex);
printf("-----------------------------------------------------\n");
#ifndef BT_NO_PROFILE
CProfileManager::Reset();
#endif //BT_NO_PROFILE
#ifndef B3_NO_PROFILE
b3ProfileManager::Reset();
#endif //B3_NO_PROFILE
window = new btgDefaultOpenGLWindow();
window = new b3gDefaultOpenGLWindow();
btgWindowConstructionInfo wci(g_OpenGLWidth,g_OpenGLHeight);
b3gWindowConstructionInfo wci(g_OpenGLWidth,g_OpenGLHeight);
window->createWindow(wci);
window->setResizeCallback(MyResizeCallback);
@@ -535,7 +535,7 @@ int main(int argc, char* argv[])
float dx=0;
if (1)
{
BT_PROFILE("font sth_draw_text");
B3_PROFILE("font sth_draw_text");
glEnable(GL_BLEND);
GLint err = glGetError();
@@ -579,7 +579,7 @@ int main(int argc, char* argv[])
window->setWheelCallback(btDefaultWheelCallback);
window->setWheelCallback(b3DefaultWheelCallback);
@@ -640,8 +640,8 @@ int main(int argc, char* argv[])
printf("-----------------------------------------------------\n");
do
{
CProfileManager::Reset();
CProfileManager::Increment_Frame_Counter();
b3ProfileManager::Reset();
b3ProfileManager::Increment_Frame_Counter();
// render.reshape(g_OpenGLWidth,g_OpenGLHeight);
@@ -654,7 +654,7 @@ int main(int argc, char* argv[])
if (!gPause)
{
BT_PROFILE("clientMoveAndDisplay");
B3_PROFILE("clientMoveAndDisplay");
demo->clientMoveAndDisplay();
}
@@ -664,16 +664,16 @@ int main(int argc, char* argv[])
}
{
BT_PROFILE("renderScene");
B3_PROFILE("renderScene");
demo->renderScene();
}
/*if (demo->getDynamicsWorld() && demo->getDynamicsWorld()->getNumCollisionObjects())
{
BT_PROFILE("renderPhysicsWorld");
b3AlignedObjectArray<btCollisionObject*> arr = demo->getDynamicsWorld()->getCollisionObjectArray();
btCollisionObject** colObjArray = &arr[0];
B3_PROFILE("renderPhysicsWorld");
b3AlignedObjectArray<b3CollisionObject*> arr = demo->getDynamicsWorld()->getCollisionObjectArray();
b3CollisionObject** colObjArray = &arr[0];
render.renderPhysicsWorld(demo->getDynamicsWorld()->getNumCollisionObjects(),colObjArray, syncOnly);
syncOnly = true;
@@ -681,20 +681,20 @@ int main(int argc, char* argv[])
}
*/
{
BT_PROFILE("gui->draw");
B3_PROFILE("gui->draw");
gui->draw(g_OpenGLWidth,g_OpenGLHeight);
}
{
BT_PROFILE("window->endRendering");
B3_PROFILE("window->endRendering");
window->endRendering();
}
{
BT_PROFILE("glFinish");
B3_PROFILE("glFinish");
}
if (dump_timings)
CProfileManager::dumpAll();
b3ProfileManager::dumpAll();
if (f)
{
@@ -715,7 +715,7 @@ int main(int argc, char* argv[])
demo->exitPhysics();
CProfileManager::CleanupMemory();
b3ProfileManager::CleanupMemory();
delete demo;
if (f)
fclose(f);

40
demo/gpudemo/premake4.lua → Demos3/GpuDemos/premake4.lua
View File

@@ -4,7 +4,7 @@ function createProject(vendor)
if (hasCL) then
project ("Bullet3_OpenCL_gpu_demo_" .. vendor)
project ("App_Bullet3_OpenCL_Demos_" .. vendor)
initOpenCL(vendor)
@@ -20,8 +20,7 @@ function createProject(vendor)
includedirs {
"..",
"../../src",
"../../btgui",
"../../opencl"
"../../btgui"
}
links {
@@ -30,21 +29,23 @@ function createProject(vendor)
"Bullet3Geometry",
"Bullet3Collision",
"Bullet3Dynamics",
"Bullet2FileLoader"
"Bullet2FileLoader",
"Bullet3OpenCL_" .. vendor
}
files {
"**.cpp",
"**.h",
"../ObjLoader/string_extra.cpp",
"../ObjLoader/string_extra.h",
"../ObjLoader/objLoader.cpp",
"../ObjLoader/objLoader.h",
"../ObjLoader/obj_parser.cpp",
"../ObjLoader/obj_parser.h",
"../ObjLoader/list.cpp",
"../ObjLoader/list.h",
"../Wavefront/string_extra.cpp",
"../Wavefront/string_extra.h",
"../Wavefront/objLoader.cpp",
"../Wavefront/objLoader.h",
"../Wavefront/obj_parser.cpp",
"../Wavefront/obj_parser.h",
"../Wavefront/list.cpp",
"../Wavefront/list.h",
"../../btgui/OpenGLWindow/GLInstancingRenderer.cpp",
@@ -60,21 +61,6 @@ function createProject(vendor)
"../../btgui/OpenGLTrueTypeFont/opengl_fontstashcallbacks.cpp",
"../../btgui/OpenGLTrueTypeFont/opengl_fontstashcallbacks.h",
"../../btgui/FontFiles/OpenSans.cpp",
"../../opencl/basic_initialize/b3OpenCLUtils.cpp",
"../../opencl/basic_initialize/b3OpenCLUtils.h",
"../../opencl/gpu_broadphase/host/b3GpuSapBroadphase.cpp",
"../../opencl/gpu_narrowphase/host/**.cpp",
"../../opencl/gpu_narrowphase/host/**.h",
"../../opencl/parallel_primitives/host/btBoundSearchCL.cpp",
"../../opencl/parallel_primitives/host/btBoundSearchCL.h",
"../../opencl/parallel_primitives/host/btFillCL.cpp",
"../../opencl/parallel_primitives/host/btFillCL.h",
"../../opencl/parallel_primitives/host/btPrefixScanCL.cpp",
"../../opencl/parallel_primitives/host/btPrefixScanCL.h",
"../../opencl/parallel_primitives/host/btRadixSort32CL.cpp",
"../../opencl/parallel_primitives/host/btRadixSort32CL.h",
"../../opencl/gpu_rigidbody/host/**.cpp",
"../../opencl/gpu_rigidbody/host/**.h",
}

4
demo/gpudemo/rigidbody/Bullet2FileDemo.cpp → Demos3/GpuDemos/rigidbody/Bullet2FileDemo.cpp
View File

@@ -15,7 +15,7 @@ Bullet2FileDemo::~Bullet2FileDemo()
void Bullet2FileDemo::setupScene(const ConstructionInfo& ci)
{
btAssert(ci.m_instancingRenderer);
b3Assert(ci.m_instancingRenderer);
const char* fileName="data/testFile.bullet";
@@ -35,7 +35,7 @@ void Bullet2FileDemo::setupScene(const ConstructionInfo& ci)
fclose(f);
createScene(*ci.m_instancingRenderer,*m_data->m_np,*m_data->m_rigidBodyPipeline,relativeFileName);
}
// m_loader = new btBulletDataExtractor(*ci.m_instancingRenderer,*m_data->m_np,*m_data->m_rigidBodyPipeline);
// m_loader = new b3BulletDataExtractor(*ci.m_instancingRenderer,*m_data->m_np,*m_data->m_rigidBodyPipeline);
// m_loader->convertAllObjects(bulletFile);
b3Vector3 pos(-20,10,0);

2
demo/gpudemo/rigidbody/Bullet2FileDemo.h → Demos3/GpuDemos/rigidbody/Bullet2FileDemo.h
View File

@@ -7,7 +7,7 @@
class Bullet2FileDemo : public GpuRigidBodyDemo
{
class btBulletDataExtractor* m_loader;
class b3BulletDataExtractor* m_loader;
public:

90
demo/gpudemo/rigidbody/BulletDataExtractor.cpp → Demos3/GpuDemos/rigidbody/BulletDataExtractor.cpp
View File

@@ -9,7 +9,7 @@ float Y_GAP = 2.f;
float Z_GAP = 2.3f;
#include "BulletDataExtractor.h"
#include "Bullet3Serialize/Bullet2FileLoader/btBulletFile.h"
#include "Bullet3Serialize/Bullet2FileLoader/b3BulletFile.h"
bool keepStaticObjects = true;
extern bool enableExperimentalCpuConcaveCollision;
@@ -19,14 +19,14 @@ extern bool enableExperimentalCpuConcaveCollision;
#include "OpenGLWindow/GLInstancingRenderer.h"
//#include "LinearMath/btQuickprof.h"
//#include "LinearMath/b3Quickprof.h"
#include "Bullet3Common/b3Quaternion.h"
#include "Bullet3Common/b3Matrix3x3.h"
#include "gpu_narrowphase/host/b3ConvexUtility.h"
#include "Bullet3OpenCL/NarrowphaseCollision/b3ConvexUtility.h"
#include "OpenGLWindow/ShapeData.h"
#include "../../ObjLoader/objLoader.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "../../Wavefront/objLoader.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
///work-in-progress
///This ReadBulletSample is kept as simple as possible without dependencies to the Bullet SDK.
@@ -54,7 +54,7 @@ struct GraphicsShape
struct InstanceGroup
{
Bullet::btCollisionShapeData* m_shape;
Bullet3SerializeBullet2::b3CollisionShapeData* m_shape;
int m_collisionShapeIndex;
b3AlignedObjectArray<bParse::bStructHandle*> m_rigidBodies;
@@ -80,7 +80,7 @@ void createScene( GLInstancingRenderer& renderer,b3GpuNarrowPhase& np, b3GpuRigi
bool verboseDumpAllTypes = false;
bParse::btBulletFile* bulletFile2 = new bParse::btBulletFile(fileName);
bParse::b3BulletFile* bulletFile2 = new bParse::b3BulletFile(fileName);
bool ok = (bulletFile2->getFlags()& bParse::FD_OK)!=0;
@@ -107,7 +107,7 @@ void createScene( GLInstancingRenderer& renderer,b3GpuNarrowPhase& np, b3GpuRigi
}
btBulletDataExtractor extractor(renderer,np,rbWorld);
b3BulletDataExtractor extractor(renderer,np,rbWorld);
extractor.convertAllObjects(bulletFile2);
delete bulletFile2;
@@ -181,24 +181,24 @@ CONCAVE_SHAPES_END_HERE,
};
btBulletDataExtractor::btBulletDataExtractor(GLInstancingRenderer& renderer, b3GpuNarrowPhase& np, b3GpuRigidBodyPipeline& rbWorld)
b3BulletDataExtractor::b3BulletDataExtractor(GLInstancingRenderer& renderer, b3GpuNarrowPhase& np, b3GpuRigidBodyPipeline& rbWorld)
:m_renderer(renderer), m_np(np), m_rbPipeline(rbWorld)
{
}
btBulletDataExtractor::~btBulletDataExtractor()
b3BulletDataExtractor::~b3BulletDataExtractor()
{
}
void btBulletDataExtractor::convertAllObjects(bParse::btBulletFile* bulletFile2)
void b3BulletDataExtractor::convertAllObjects(bParse::b3BulletFile* bulletFile2)
{
int i;
for (i=0;i<bulletFile2->m_collisionShapes.size();i++)
{
Bullet::btCollisionShapeData* shapeData = (Bullet::btCollisionShapeData*)bulletFile2->m_collisionShapes[i];
Bullet3SerializeBullet2::b3CollisionShapeData* shapeData = (Bullet3SerializeBullet2::b3CollisionShapeData*)bulletFile2->m_collisionShapes[i];
if (shapeData->m_name)
printf("converting shape %s\n", shapeData->m_name);
int shapeIndex = convertCollisionShape(shapeData);
@@ -216,8 +216,8 @@ void btBulletDataExtractor::convertAllObjects(bParse::btBulletFile* bulletFile2)
for (i=0;i<bulletFile2->m_rigidBodies.size();i++)
{
Bullet::btRigidBodyFloatData* colObjData = (Bullet::btRigidBodyFloatData*)bulletFile2->m_rigidBodies[i];
Bullet::btCollisionShapeData* shapeData = (Bullet::btCollisionShapeData*)colObjData->m_collisionObjectData.m_collisionShape;
Bullet3SerializeBullet2::b3RigidBodyFloatData* colObjData = (Bullet3SerializeBullet2::b3RigidBodyFloatData*)bulletFile2->m_rigidBodies[i];
Bullet3SerializeBullet2::b3CollisionShapeData* shapeData = (Bullet3SerializeBullet2::b3CollisionShapeData*)colObjData->m_collisionObjectData.m_collisionShape;
for (int j=0;j<m_instanceGroups.size();j++)
{
if (m_instanceGroups[j]->m_shape == shapeData)
@@ -237,10 +237,10 @@ void btBulletDataExtractor::convertAllObjects(bParse::btBulletFile* bulletFile2)
for (int j=0;j<m_instanceGroups[i]->m_rigidBodies.size();j++)
{
Bullet::btRigidBodyFloatData* colObjData = (Bullet::btRigidBodyFloatData*)m_instanceGroups[i]->m_rigidBodies[j];
Bullet3SerializeBullet2::b3RigidBodyFloatData* colObjData = (Bullet3SerializeBullet2::b3RigidBodyFloatData*)m_instanceGroups[i]->m_rigidBodies[j];
b3Matrix3x3 mat;
mat.deSerializeFloat((const btMatrix3x3FloatData&)colObjData->m_collisionObjectData.m_worldTransform.m_basis);
mat.deSerializeFloat((const b3Matrix3x3FloatData&)colObjData->m_collisionObjectData.m_worldTransform.m_basis);
b3Quaternion orn;
mat.getRotation(orn);
float quaternion[4] = {orn[0],orn[1],orn[2],orn[3]};
@@ -283,7 +283,7 @@ void btBulletDataExtractor::convertAllObjects(bParse::btBulletFile* bulletFile2)
int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData* shapeData )
int b3BulletDataExtractor::convertCollisionShape( Bullet3SerializeBullet2::b3CollisionShapeData* shapeData )
{
int shapeIndex = -1;
@@ -291,7 +291,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
{
case STATIC_PLANE_PROXYTYPE:
{
Bullet::btStaticPlaneShapeData* planeData = (Bullet::btStaticPlaneShapeData*)shapeData;
Bullet3SerializeBullet2::b3StaticPlaneShapeData* planeData = (Bullet3SerializeBullet2::b3StaticPlaneShapeData*)shapeData;
shapeIndex = createPlaneShape(planeData->m_planeNormal,planeData->m_planeConstant, planeData->m_localScaling);
break;
}
@@ -303,7 +303,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
case MULTI_SPHERE_SHAPE_PROXYTYPE:
case CONVEX_HULL_SHAPE_PROXYTYPE:
{
Bullet::btConvexInternalShapeData* bsd = (Bullet::btConvexInternalShapeData*)shapeData;
Bullet3SerializeBullet2::b3ConvexInternalShapeData* bsd = (Bullet3SerializeBullet2::b3ConvexInternalShapeData*)shapeData;
switch (shapeData->m_shapeType)
{
@@ -319,10 +319,10 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
}
case CONVEX_HULL_SHAPE_PROXYTYPE:
{
Bullet::btConvexHullShapeData* convexData = (Bullet::btConvexHullShapeData*)bsd;
Bullet3SerializeBullet2::b3ConvexHullShapeData* convexData = (Bullet3SerializeBullet2::b3ConvexHullShapeData*)bsd;
int numPoints = convexData->m_numUnscaledPoints;
b3Vector3 localScaling;
localScaling.deSerializeFloat((btVector3FloatData&)bsd->m_localScaling);
localScaling.deSerializeFloat((b3Vector3FloatData&)bsd->m_localScaling);
b3AlignedObjectArray<b3Vector3> tmpPoints;
int i;
if (convexData->m_unscaledPointsFloatPtr)
@@ -355,7 +355,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
#if 0
case CAPSULE_SHAPE_PROXYTYPE:
{
btCapsuleShapeData* capData = (btCapsuleShapeData*)shapeData;
b3CapsuleShapeData* capData = (b3CapsuleShapeData*)shapeData;
switch (capData->m_upAxis)
{
case 0:
@@ -375,7 +375,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
}
default:
{
printf("error: wrong up axis for btCapsuleShape\n");
printf("error: wrong up axis for b3CapsuleShape\n");
}
};
@@ -384,7 +384,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
}
case CYLINDER_SHAPE_PROXYTYPE:
{
btCylinderShapeData* cylData = (btCylinderShapeData*) shapeData;
b3CylinderShapeData* cylData = (b3CylinderShapeData*) shapeData;
b3Vector3 halfExtents = implicitShapeDimensions+margin;
switch (cylData->m_upAxis)
{
@@ -416,7 +416,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
}
case MULTI_SPHERE_SHAPE_PROXYTYPE:
{
btMultiSphereShapeData* mss = (btMultiSphereShapeData*)bsd;
b3MultiSphereShapeData* mss = (b3MultiSphereShapeData*)bsd;
int numSpheres = mss->m_localPositionArraySize;
int i;
for ( i=0;i<numSpheres;i++)
@@ -424,7 +424,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
tmpPos[i].deSerializeFloat(mss->m_localPositionArrayPtr[i].m_pos);
radii[i] = mss->m_localPositionArrayPtr[i].m_radius;
}
shape = new btMultiSphereShape(&tmpPos[0],&radii[0],numSpheres);
shape = new b3MultiSphereShape(&tmpPos[0],&radii[0],numSpheres);
break;
}
@@ -441,13 +441,13 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
case TRIANGLE_MESH_SHAPE_PROXYTYPE:
{
Bullet::btTriangleMeshShapeData* trimesh = (Bullet::btTriangleMeshShapeData*)shapeData;
Bullet3SerializeBullet2::b3TriangleMeshShapeData* trimesh = (Bullet3SerializeBullet2::b3TriangleMeshShapeData*)shapeData;
printf("numparts = %d\n",trimesh->m_meshInterface.m_numMeshParts);
if (trimesh->m_meshInterface.m_numMeshParts)
{
for (int i=0;i<trimesh->m_meshInterface.m_numMeshParts;i++)
{
Bullet::btMeshPartData& dat = trimesh->m_meshInterface.m_meshPartsPtr[i];
Bullet3SerializeBullet2::b3MeshPartData& dat = trimesh->m_meshInterface.m_meshPartsPtr[i];
printf("numtris = %d, numverts = %d\n", dat.m_numTriangles,dat.m_numVertices);//,dat.m_vertices3f,dat.m_3indices16
printf("scaling = %f,%f,%f\n", trimesh->m_meshInterface.m_scaling.m_floats[0],trimesh->m_meshInterface.m_scaling.m_floats[1],trimesh->m_meshInterface.m_scaling.m_floats[2]);
// dat.
@@ -459,7 +459,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
//trimesh->m_meshInterface.m_scaling
}
//trimesh->m_meshInterface
//btTriangleIndexVertexArray* meshInterface = createMeshInterface(trimesh->m_meshInterface);
//b3TriangleIndexVertexArray* meshInterface = createMeshInterface(trimesh->m_meshInterface);
//scaling
@@ -473,17 +473,17 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
#if 0
case COMPOUND_SHAPE_PROXYTYPE:
{
btCompoundShapeData* compoundData = (btCompoundShapeData*)shapeData;
btCompoundShape* compoundShape = createCompoundShape();
b3CompoundShapeData* compoundData = (b3CompoundShapeData*)shapeData;
b3CompoundShape* compoundShape = createCompoundShape();
b3AlignedObjectArray<btCollisionShape*> childShapes;
b3AlignedObjectArray<b3CollisionShape*> childShapes;
for (int i=0;i<compoundData->m_numChildShapes;i++)
{
btCollisionShape* childShape = convertCollisionShape(compoundData->m_childShapePtr[i].m_childShape);
b3CollisionShape* childShape = convertCollisionShape(compoundData->m_childShapePtr[i].m_childShape);
if (childShape)
{
btTransform localTransform;
b3Transform localTransform;
localTransform.deSerializeFloat(compoundData->m_childShapePtr[i].m_transform);
compoundShape->addChildShape(localTransform,childShape);
} else
@@ -499,11 +499,11 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
case GIMPACT_SHAPE_PROXYTYPE:
{
btGImpactMeshShapeData* gimpactData = (btGImpactMeshShapeData*) shapeData;
b3GImpactMeshShapeData* gimpactData = (b3GImpactMeshShapeData*) shapeData;
if (gimpactData->m_gimpactSubType == CONST_GIMPACT_TRIMESH_SHAPE)
{
btTriangleIndexVertexArray* meshInterface = createMeshInterface(gimpactData->m_meshInterface);
btGImpactMeshShape* gimpactShape = createGimpactShape(meshInterface);
b3TriangleIndexVertexArray* meshInterface = createMeshInterface(gimpactData->m_meshInterface);
b3GImpactMeshShape* gimpactShape = createGimpactShape(meshInterface);
b3Vector3 localScaling;
localScaling.deSerializeFloat(gimpactData->m_localScaling);
gimpactShape->setLocalScaling(localScaling);
@@ -531,7 +531,7 @@ int btBulletDataExtractor::convertCollisionShape( Bullet::btCollisionShapeData*
}
int btBulletDataExtractor::createBoxShape( const Bullet::btVector3FloatData& halfDimensions, const Bullet::btVector3FloatData& localScaling, float collisionMargin)
int b3BulletDataExtractor::createBoxShape( const Bullet3SerializeBullet2::b3Vector3FloatData& halfDimensions, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling, float collisionMargin)
{
float cubeScaling[4] = {
halfDimensions.m_floats[0]*localScaling.m_floats[0]+collisionMargin,
@@ -561,14 +561,14 @@ int btBulletDataExtractor::createBoxShape( const Bullet::btVector3FloatData& hal
return cubeCollisionShapeIndex;
}
int btBulletDataExtractor::createSphereShape( float radius, const Bullet::btVector3FloatData& localScaling, float collisionMargin)
int b3BulletDataExtractor::createSphereShape( float radius, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling, float collisionMargin)
{
printf("createSphereShape with radius %f\n",radius);
return -1;
}
int btBulletDataExtractor::createPlaneShape( const Bullet::btVector3FloatData& planeNormal, float planeConstant, const Bullet::btVector3FloatData& localScaling)
int b3BulletDataExtractor::createPlaneShape( const Bullet3SerializeBullet2::b3Vector3FloatData& planeNormal, float planeConstant, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling)
{
printf("createPlaneShape with normal %f,%f,%f and planeConstant\n",planeNormal.m_floats[0], planeNormal.m_floats[1],planeNormal.m_floats[2],planeConstant);
return -1;
@@ -576,7 +576,7 @@ int btBulletDataExtractor::createPlaneShape( const Bullet::btVector3FloatData& p
GraphicsShape* btBulletDataExtractor::createGraphicsShapeFromWavefrontObj(objLoader* obj)
GraphicsShape* b3BulletDataExtractor::createGraphicsShapeFromWavefrontObj(objLoader* obj)
{
b3AlignedObjectArray<GraphicsVertex>* vertices = new b3AlignedObjectArray<GraphicsVertex>;
{
@@ -590,7 +590,7 @@ GraphicsShape* btBulletDataExtractor::createGraphicsShapeFromWavefrontObj(objLoa
vtx.xyzw[1] = obj->vertexList[v]->e[1];
vtx.xyzw[2] = obj->vertexList[v]->e[2];
b3Vector3 n(vtx.xyzw[0],vtx.xyzw[1],vtx.xyzw[2]);
if (n.length2()>SIMD_EPSILON)
if (n.length2()>B3_EPSILON)
{
n.normalize();
vtx.normal[0] = n[0];
@@ -701,7 +701,7 @@ GraphicsShape* btBulletDataExtractor::createGraphicsShapeFromWavefrontObj(objLoa
}
GraphicsShape* btBulletDataExtractor::createGraphicsShapeFromConvexHull(const b3Vector3* tmpPoints, int numPoints)
GraphicsShape* b3BulletDataExtractor::createGraphicsShapeFromConvexHull(const b3Vector3* tmpPoints, int numPoints)
{
b3ConvexUtility* utilPtr = new b3ConvexUtility();
bool merge = true;
@@ -714,7 +714,7 @@ GraphicsShape* btBulletDataExtractor::createGraphicsShapeFromConvexHull(const b3
b3AlignedObjectArray<int>* indicesPtr = new b3AlignedObjectArray<int>;
for (int f=0;f<utilPtr->m_faces.size();f++)
{
const btMyFace& face = utilPtr->m_faces[f];
const b3MyFace& face = utilPtr->m_faces[f];
b3Vector3 normal(face.m_plane[0],face.m_plane[1],face.m_plane[2]);
if (face.m_indices.size()>2)
{

18
demo/gpudemo/rigidbody/BulletDataExtractor.h → Demos3/GpuDemos/rigidbody/BulletDataExtractor.h
View File

@@ -14,10 +14,10 @@ void createScene(GLInstancingRenderer& renderer,b3GpuNarrowPhase& np, b3GpuRigid
namespace bParse
{
class btBulletFile;
class b3BulletFile;
};
class btBulletDataExtractor
class b3BulletDataExtractor
{
GLInstancingRenderer& m_renderer;
b3GpuNarrowPhase& m_np;
@@ -28,20 +28,20 @@ class btBulletDataExtractor
b3AlignedObjectArray<struct InstanceGroup*> m_instanceGroups;
b3AlignedObjectArray<struct GraphicsShape*> m_graphicsShapes;
btBulletDataExtractor(GLInstancingRenderer& renderer, b3GpuNarrowPhase& np, b3GpuRigidBodyPipeline& rbPipeline);
b3BulletDataExtractor(GLInstancingRenderer& renderer, b3GpuNarrowPhase& np, b3GpuRigidBodyPipeline& rbPipeline);
virtual ~btBulletDataExtractor();
virtual ~b3BulletDataExtractor();
virtual void convertAllObjects(bParse::btBulletFile* bulletFile);
virtual void convertAllObjects(bParse::b3BulletFile* bulletFile);
//return -1 for invalid
virtual int convertCollisionShape( Bullet::btCollisionShapeData* shapeData );
virtual int convertCollisionShape( Bullet3SerializeBullet2::b3CollisionShapeData* shapeData );
virtual int createPlaneShape( const Bullet::btVector3FloatData& planeNormal, float planeConstant, const Bullet::btVector3FloatData& localScaling);
virtual int createPlaneShape( const Bullet3SerializeBullet2::b3Vector3FloatData& planeNormal, float planeConstant, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling);
virtual int createBoxShape( const Bullet::btVector3FloatData& halfDimensions, const Bullet::btVector3FloatData& localScaling, float collisionMargin);
virtual int createBoxShape( const Bullet3SerializeBullet2::b3Vector3FloatData& halfDimensions, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling, float collisionMargin);
virtual int createSphereShape( float radius, const Bullet::btVector3FloatData& localScaling, float collisionMargin);
virtual int createSphereShape( float radius, const Bullet3SerializeBullet2::b3Vector3FloatData& localScaling, float collisionMargin);
static GraphicsShape* createGraphicsShapeFromConvexHull(const b3Vector3* tmpPoints, int numPoints);
static GraphicsShape* createGraphicsShapeFromWavefrontObj(class objLoader* obj);

98
demo/gpudemo/rigidbody/ConcaveScene.cpp → Demos3/GpuDemos/rigidbody/ConcaveScene.cpp
View File

@@ -5,18 +5,18 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "gpu_rigidbody/host/b3Config.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
#include "Bullet3OpenCL/RigidBody/b3Config.h"
#include "GpuRigidBodyDemoInternalData.h"
#include"../../ObjLoader/objLoader.h"
#include"../../Wavefront/objLoader.h"
#include "Bullet3Common/b3Transform.h"
#include "OpenGLWindow/GLInstanceGraphicsShape.h"
@@ -39,7 +39,7 @@ GLInstanceGraphicsShape* createGraphicsShapeFromWavefrontObj(objLoader* obj)
vtx.xyzw[1] = obj->vertexList[v]->e[1];
vtx.xyzw[2] = obj->vertexList[v]->e[2];
b3Vector3 n(vtx.xyzw[0],vtx.xyzw[1],vtx.xyzw[2]);
if (n.length2()>SIMD_EPSILON)
if (n.length2()>B3_EPSILON)
{
n.normalize();
vtx.normal[0] = n[0];
@@ -216,7 +216,7 @@ void ConcaveScene::createConcaveMesh(const ConstructionInfo& ci, const char* fil
int shapeId = ci.m_instancingRenderer->registerShape(&shape->m_vertices->at(0).xyzw[0], shape->m_numvertices, &shape->m_indices->at(0), shape->m_numIndices);
b3Quaternion orn(0,0,0,1);
btVector4 color(0.3,0.3,1,1.f);//0.5);//1.f
b3Vector4 color(0.3,0.3,1,1.f);//0.5);//1.f
{
@@ -256,7 +256,7 @@ void ConcaveScene::setupScene(const ConstructionInfo& ci)
b3Vector3 shift1(0,-50,0);//0,230,80);//150,-100,-120);
btVector4 scaling(4,4,4,1);
b3Vector4 scaling(4,4,4,1);
// createConcaveMesh(ci,"data/plane100.obj",shift1,scaling);
//createConcaveMesh(ci,"data/plane100.obj",shift,scaling);
@@ -278,12 +278,12 @@ void ConcaveScene::setupScene(const ConstructionInfo& ci)
int mask=1;
int index=0;
{
btVector4 scaling(400,0.001,400,1);
b3Vector4 scaling(400,0.001,400,1);
int colIndex = m_data->m_np->registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
b3Vector3 position(0,-2,0);
b3Quaternion orn(0,0,0,1);
btVector4 color(0,0,1,1);
b3Vector4 color(0,0,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(0.f,position,orn,colIndex,index,false);
@@ -320,15 +320,15 @@ void ConcaveScene::createDynamicObjects(const ConstructionInfo& ci)
if (1)
{
int curColor = 0;
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,1,1,1),
btVector4(1,1,0.3,1),
btVector4(0.3,1,1,1),
btVector4(0.3,0.3,1,1),
b3Vector4(1,1,1,1),
b3Vector4(1,1,0.3,1),
b3Vector4(0.3,1,1,1),
b3Vector4(0.3,0.3,1,1),
};
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
int colIndex = m_data->m_np->registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
for (int i=0;i<ci.arraySizeX;i++)
{
@@ -342,7 +342,7 @@ void ConcaveScene::createDynamicObjects(const ConstructionInfo& ci)
b3Vector3 position(-(ci.arraySizeX/2)*CONCAVE_GAPX+i*CONCAVE_GAPX,150+j*CONCAVE_GAPY,-(ci.arraySizeZ/2)*CONCAVE_GAPZ+k*CONCAVE_GAPZ);
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
@@ -379,7 +379,7 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
b3Vector3 shift(0,0,0);//0,230,80);//150,-100,-120);
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
FILE* f = 0;
char relativeFileName[1024];
@@ -427,12 +427,12 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
b3AlignedObjectArray<GLInstanceVertex> vertices;
int stride2 = sizeof(GLInstanceVertex);
btAssert(stride2 == strideInBytes);
b3Assert(stride2 == strideInBytes);
{
b3AlignedObjectArray<btGpuChildShape> childShapes;
b3AlignedObjectArray<b3GpuChildShape> childShapes;
int numChildShapes = objData->objectCount;
@@ -448,7 +448,7 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
//for now, only support polyhedral child shapes
btGpuChildShape child;
b3GpuChildShape child;
b3Vector3 pos(0,0,0);
b3Quaternion orn(0,0,0,1);
@@ -476,7 +476,7 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
}
} else
{
btAssert(0);
b3Assert(0);
}
}
@@ -545,12 +545,12 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
//int shapeId = ci.m_instancingRenderer->registerShape(&cube_vertices[0],numVertices,cube_indices,numIndices);
int shapeId = ci.m_instancingRenderer->registerShape(&vertexArray[0].xyzw[0],vertexArray.size(),&indexArray[0],indexArray.size());
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,0,1,1),
btVector4(0,1,1,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,0,1,1),
b3Vector4(0,1,1,1),
};
int curColor = 0;
@@ -569,10 +569,10 @@ void ConcaveCompound2Scene::createDynamicObjects(const ConstructionInfo& ci)
// b3Quaternion orn(0,0,0,1);
b3Quaternion orn(b3Vector3(0,0,1),1.8);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);
@@ -605,7 +605,7 @@ void ConcaveCompoundScene::createDynamicObjects(const ConstructionInfo& ci)
GLInstanceVertex* cubeVerts = (GLInstanceVertex*)&cube_vertices[0];
int stride2 = sizeof(GLInstanceVertex);
btAssert(stride2 == strideInBytes);
b3Assert(stride2 == strideInBytes);
{
int childColIndex = m_data->m_np->registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
@@ -617,12 +617,12 @@ b3Vector3 childPositions[3] = {
b3Vector3(0,0,2)
};
b3AlignedObjectArray<btGpuChildShape> childShapes;
b3AlignedObjectArray<b3GpuChildShape> childShapes;
int numChildShapes = 3;
for (int i=0;i<numChildShapes;i++)
{
//for now, only support polyhedral child shapes
btGpuChildShape child;
b3GpuChildShape child;
child.m_shapeIndex = childColIndex;
b3Vector3 pos = childPositions[i];
b3Quaternion orn(0,0,0,1);
@@ -662,12 +662,12 @@ b3Vector3 childPositions[3] = {
//int shapeId = ci.m_instancingRenderer->registerShape(&cube_vertices[0],numVertices,cube_indices,numIndices);
int shapeId = ci.m_instancingRenderer->registerShape(&vertexArray[0].xyzw[0],vertexArray.size(),&indexArray[0],indexArray.size());
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,0,1,1),
btVector4(0,1,1,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,0,1,1),
b3Vector4(0,1,1,1),
};
int curColor = 0;
@@ -684,10 +684,10 @@ b3Vector3 childPositions[3] = {
//b3Quaternion orn(0,0,0,1);
b3Quaternion orn(b3Vector3(1,0,0),0.7);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);
@@ -713,12 +713,12 @@ void ConcaveSphereScene::setupScene(const ConstructionInfo& ci)
void ConcaveSphereScene::createDynamicObjects(const ConstructionInfo& ci)
{
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,1,1,1),
btVector4(1,1,0,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,1,1,1),
b3Vector4(1,1,0,1),
};
int index=0;
@@ -743,10 +743,10 @@ void ConcaveSphereScene::createDynamicObjects(const ConstructionInfo& ci)
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(radius,radius,radius,1);
b3Vector4 scaling(radius,radius,radius,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(prevGraphicsShapeIndex,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);

0
demo/gpudemo/rigidbody/ConcaveScene.h → Demos3/GpuDemos/rigidbody/ConcaveScene.h
View File

54
demo/gpudemo/rigidbody/GpuCompoundScene.cpp → Demos3/GpuDemos/rigidbody/GpuCompoundScene.cpp
View File

@@ -5,16 +5,16 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "gpu_rigidbody/host/b3Config.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
#include "Bullet3OpenCL/RigidBody/b3Config.h"
#include "GpuRigidBodyDemoInternalData.h"
#include "Bullet3Common/b3Transform.h"
@@ -34,7 +34,7 @@ void GpuCompoundScene::setupScene(const ConstructionInfo& ci)
GLInstanceVertex* cubeVerts = (GLInstanceVertex*)&cube_vertices[0];
int stride2 = sizeof(GLInstanceVertex);
btAssert(stride2 == strideInBytes);
b3Assert(stride2 == strideInBytes);
int index=0;
int colIndex = -1;
b3AlignedObjectArray<GLInstanceVertex> vertexArray;
@@ -50,12 +50,12 @@ void GpuCompoundScene::setupScene(const ConstructionInfo& ci)
};
b3AlignedObjectArray<btGpuChildShape> childShapes;
b3AlignedObjectArray<b3GpuChildShape> childShapes;
int numChildShapes = 3;
for (int i=0;i<numChildShapes;i++)
{
//for now, only support polyhedral child shapes
btGpuChildShape child;
b3GpuChildShape child;
child.m_shapeIndex = childColIndex;
b3Vector3 pos = childPositions[i];
b3Quaternion orn(0,0,0,1);
@@ -96,12 +96,12 @@ void GpuCompoundScene::setupScene(const ConstructionInfo& ci)
//int shapeId = ci.m_instancingRenderer->registerShape(&cube_vertices[0],numVertices,cube_indices,numIndices);
int shapeId = ci.m_instancingRenderer->registerShape(&vertexArray[0].xyzw[0],vertexArray.size(),&indexArray[0],indexArray.size());
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,0,1,1),
btVector4(0,1,1,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,0,1,1),
b3Vector4(0,1,1,1),
};
int curColor = 0;
@@ -117,10 +117,10 @@ void GpuCompoundScene::setupScene(const ConstructionInfo& ci)
//b3Quaternion orn(0,0,0,1);
b3Quaternion orn(b3Vector3(1,0,0),0.7);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);
@@ -179,7 +179,7 @@ void GpuCompoundScene::createStaticEnvironment(const ConstructionInfo& ci)
{
int numVertices = sizeof(point_sphere_vertices)/strideInBytes;
int numIndices = sizeof(point_sphere_indices)/sizeof(int);
prevGraphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,BT_GL_POINTS);
prevGraphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,B3_GL_POINTS);
} else
{
if (radius>=10)
@@ -197,12 +197,12 @@ void GpuCompoundScene::createStaticEnvironment(const ConstructionInfo& ci)
}
}
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,1,1,1),
btVector4(1,1,0,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,1,1,1),
b3Vector4(1,1,0,1),
};
int curColor = 1;
@@ -217,10 +217,10 @@ void GpuCompoundScene::createStaticEnvironment(const ConstructionInfo& ci)
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(radius,radius,radius,1);
b3Vector4 scaling(radius,radius,radius,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(prevGraphicsShapeIndex,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);
@@ -241,8 +241,8 @@ void GpuCompoundPlaneScene::createStaticEnvironment(const ConstructionInfo& ci)
b3Vector3 position(0,0,0);
b3Quaternion orn(0,0,0,1);
// b3Quaternion orn(b3Vector3(1,0,0),0.3);
btVector4 color(0,0,1,1);
btVector4 scaling(100,0.01,100,1);
b3Vector4 color(0,0,1,1);
b3Vector4 scaling(100,0.01,100,1);
int strideInBytes = 9*sizeof(float);
int numVertices = sizeof(cube_vertices)/strideInBytes;
int numIndices = sizeof(cube_indices)/sizeof(int);

0
demo/gpudemo/rigidbody/GpuCompoundScene.h → Demos3/GpuDemos/rigidbody/GpuCompoundScene.h
View File

36
demo/gpudemo/rigidbody/GpuConvexScene.cpp → Demos3/GpuDemos/rigidbody/GpuConvexScene.cpp
View File

@@ -5,16 +5,16 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "gpu_rigidbody/host/b3Config.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
#include "Bullet3OpenCL/RigidBody/b3Config.h"
#include "GpuRigidBodyDemoInternalData.h"
#include "../gwenUserInterface.h"
#include "Bullet3Dynamics/ConstraintSolver/b3Point2PointConstraint.h"
@@ -73,12 +73,12 @@ int GpuConvexScene::createDynamicsObjects2(const ConstructionInfo& ci, const flo
{
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,1,1,1),
btVector4(1,1,0,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,1,1,1),
b3Vector4(1,1,0,1),
};
int curColor = 0;
@@ -104,10 +104,10 @@ int GpuConvexScene::createDynamicsObjects2(const ConstructionInfo& ci, const flo
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(1,1,1,1);
b3Vector4 scaling(1,1,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index,false);
@@ -140,12 +140,12 @@ void GpuConvexScene::createStaticEnvironment(const ConstructionInfo& ci)
{
btVector4 scaling(400,1,400,1);
b3Vector4 scaling(400,1,400,1);
int colIndex = m_data->m_np->registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
b3Vector3 position(0,0,0);
b3Quaternion orn(0,0,0,1);
btVector4 color(0,0,1,1);
b3Vector4 color(0,0,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(0.f,position,orn,colIndex,index,false);
@@ -162,8 +162,8 @@ void GpuConvexPlaneScene::createStaticEnvironment(const ConstructionInfo& ci)
b3Vector3 position(0,0,0);
b3Quaternion orn(0,0,0,1);
// b3Quaternion orn(b3Vector3(1,0,0),0.3);
btVector4 color(0,0,1,1);
btVector4 scaling(100,0.001,100,1);
b3Vector4 color(0,0,1,1);
b3Vector4 scaling(100,0.001,100,1);
int strideInBytes = 9*sizeof(float);
int numVertices = sizeof(cube_vertices)/strideInBytes;
int numIndices = sizeof(cube_indices)/sizeof(int);

0
demo/gpudemo/rigidbody/GpuConvexScene.h → Demos3/GpuDemos/rigidbody/GpuConvexScene.h
View File

40
demo/gpudemo/rigidbody/GpuRigidBodyDemo.cpp → Demos3/GpuDemos/rigidbody/GpuRigidBodyDemo.cpp
View File

@@ -3,20 +3,20 @@
#include "OpenGLWindow/ShapeData.h"
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "gpu_rigidbody/host/b3Config.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
#include "Bullet3OpenCL/RigidBody/b3Config.h"
#include "GpuRigidBodyDemoInternalData.h"
#include "Bullet3Collision/BroadPhaseCollision/b3DynamicBvhBroadphase.h"
static btKeyboardCallback oldCallback = 0;
static b3KeyboardCallback oldCallback = 0;
extern bool gReset;
#define MSTRINGIFY(A) #A
@@ -76,7 +76,7 @@ static void PairKeyboardCallback(int key, int state)
gReset = true;
}
//btDefaultKeyboardCallback(key,state);
//b3DefaultKeyboardCallback(key,state);
oldCallback(key,state);
}
@@ -156,21 +156,21 @@ void GpuRigidBodyDemo::clientMoveAndDisplay()
bool animate=true;
int numObjects= m_data->m_rigidBodyPipeline->getNumBodies();
//m_instancingRenderer->getInternalData()->m_totalNumInstances;
btVector4* positions = 0;
b3Vector4* positions = 0;
if (animate && numObjects)
{
BT_PROFILE("gl2cl");
B3_PROFILE("gl2cl");
if (!m_data->m_instancePosOrnColor)
{
GLuint vbo = m_instancingRenderer->getInternalData()->m_vbo;
int arraySizeInBytes = numObjects * (3)*sizeof(btVector4);
int arraySizeInBytes = numObjects * (3)*sizeof(b3Vector4);
glBindBuffer(GL_ARRAY_BUFFER, vbo);
cl_bool blocking= CL_TRUE;
positions= (btVector4*)glMapBufferRange( GL_ARRAY_BUFFER,m_instancingRenderer->getMaxShapeCapacity(),arraySizeInBytes, GL_MAP_READ_BIT );//GL_READ_WRITE);//GL_WRITE_ONLY
positions= (b3Vector4*)glMapBufferRange( GL_ARRAY_BUFFER,m_instancingRenderer->getMaxShapeCapacity(),arraySizeInBytes, GL_MAP_READ_BIT );//GL_READ_WRITE);//GL_WRITE_ONLY
GLint err = glGetError();
assert(err==GL_NO_ERROR);
m_data->m_instancePosOrnColor = new btOpenCLArray<btVector4>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_instancePosOrnColor = new b3OpenCLArray<b3Vector4>(m_clData->m_clContext,m_clData->m_clQueue);
m_data->m_instancePosOrnColor->resize(3*numObjects);
m_data->m_instancePosOrnColor->copyFromHostPointer(positions,3*numObjects,0);
glUnmapBuffer( GL_ARRAY_BUFFER);
@@ -180,16 +180,16 @@ void GpuRigidBodyDemo::clientMoveAndDisplay()
}
{
BT_PROFILE("stepSimulation");
B3_PROFILE("stepSimulation");
m_data->m_rigidBodyPipeline->stepSimulation(1./60.f);
}
if (numObjects)
{
BT_PROFILE("cl2gl_convert");
B3_PROFILE("cl2gl_convert");
int ciErrNum = 0;
cl_mem bodies = m_data->m_rigidBodyPipeline->getBodyBuffer();
btLauncherCL launch(m_clData->m_clQueue,m_data->m_copyTransformsToVBOKernel);
b3LauncherCL launch(m_clData->m_clQueue,m_data->m_copyTransformsToVBOKernel);
launch.setBuffer(bodies);
launch.setBuffer(m_data->m_instancePosOrnColor->getBufferCL());
launch.setConst(numObjects);
@@ -199,14 +199,14 @@ void GpuRigidBodyDemo::clientMoveAndDisplay()
if (animate && numObjects)
{
BT_PROFILE("cl2gl_upload");
B3_PROFILE("cl2gl_upload");
GLint err = glGetError();
assert(err==GL_NO_ERROR);
GLuint vbo = m_instancingRenderer->getInternalData()->m_vbo;
int arraySizeInBytes = numObjects * (3)*sizeof(btVector4);
int arraySizeInBytes = numObjects * (3)*sizeof(b3Vector4);
glBindBuffer(GL_ARRAY_BUFFER, vbo);
cl_bool blocking= CL_TRUE;
positions= (btVector4*)glMapBufferRange( GL_ARRAY_BUFFER,m_instancingRenderer->getMaxShapeCapacity(),arraySizeInBytes, GL_MAP_WRITE_BIT );//GL_READ_WRITE);//GL_WRITE_ONLY
positions= (b3Vector4*)glMapBufferRange( GL_ARRAY_BUFFER,m_instancingRenderer->getMaxShapeCapacity(),arraySizeInBytes, GL_MAP_WRITE_BIT );//GL_READ_WRITE);//GL_WRITE_ONLY
err = glGetError();
assert(err==GL_NO_ERROR);
m_data->m_instancePosOrnColor->copyToHostPointer(positions,3*numObjects,0);

2
demo/gpudemo/rigidbody/GpuRigidBodyDemo.h → Demos3/GpuDemos/rigidbody/GpuRigidBodyDemo.h
View File

@@ -7,7 +7,7 @@ class GpuRigidBodyDemo : public GpuDemo
{
protected:
class GLInstancingRenderer* m_instancingRenderer;
class btgWindowInterface* m_window;
class b3gWindowInterface* m_window;
struct GpuRigidBodyDemoInternalData* m_data;

6
demo/gpudemo/rigidbody/GpuRigidBodyDemoInternalData.h → Demos3/GpuDemos/rigidbody/GpuRigidBodyDemoInternalData.h
View File

@@ -1,8 +1,8 @@
#ifndef GPU_RIGIDBODY_INTERNAL_DATA_H
#define GPU_RIGIDBODY_INTERNAL_DATA_H
#include "basic_initialize/b3OpenCLUtils.h"
#include "parallel_primitives/host/btOpenCLArray.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3OpenCLArray.h"
#include "Bullet3Common/b3Vector3.h"
struct GpuRigidBodyDemoInternalData
@@ -10,7 +10,7 @@ struct GpuRigidBodyDemoInternalData
cl_kernel m_copyTransformsToVBOKernel;
btOpenCLArray<btVector4>* m_instancePosOrnColor;
b3OpenCLArray<b3Vector4>* m_instancePosOrnColor;
class b3GpuRigidBodyPipeline* m_rigidBodyPipeline;

51
demo/gpudemo/rigidbody/GpuSphereScene.cpp → Demos3/GpuDemos/rigidbody/GpuSphereScene.cpp
View File

@@ -2,19 +2,18 @@
#include "GpuRigidBodyDemo.h"
#include "Bullet3Common/b3Quickprof.h"
#include "OpenGLWindow/ShapeData.h"
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "Bullet3Common/b3Quaternion.h"
#include "OpenGLWindow/btgWindowInterface.h"
#include "gpu_broadphase/host/b3GpuSapBroadphase.h"
#include "OpenGLWindow/b3gWindowInterface.h"
#include "Bullet3OpenCL/BroadphaseCollision/b3GpuSapBroadphase.h"
#include "../GpuDemoInternalData.h"
#include "basic_initialize/b3OpenCLUtils.h"
#include "Bullet3OpenCL/Initialize/b3OpenCLUtils.h"
#include "OpenGLWindow/OpenGLInclude.h"
#include "OpenGLWindow/GLInstanceRendererInternalData.h"
#include "parallel_primitives/host/btLauncherCL.h"
#include "gpu_rigidbody/host/b3GpuRigidBodyPipeline.h"
#include "gpu_rigidbody/host/b3GpuNarrowPhase.h"
#include "gpu_rigidbody/host/b3Config.h"
#include "Bullet3OpenCL/ParallelPrimitives/b3LauncherCL.h"
#include "Bullet3OpenCL/RigidBody/b3GpuRigidBodyPipeline.h"
#include "Bullet3OpenCL/RigidBody/b3GpuNarrowPhase.h"
#include "Bullet3OpenCL/RigidBody/b3Config.h"
#include "GpuRigidBodyDemoInternalData.h"
#include "../gwenUserInterface.h"
@@ -53,7 +52,7 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
{
int numVertices = sizeof(point_sphere_vertices)/strideInBytes;
int numIndices = sizeof(point_sphere_indices)/sizeof(int);
prevGraphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,BT_GL_POINTS);
prevGraphicsShapeIndex = ci.m_instancingRenderer->registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,B3_GL_POINTS);
} else
{
if (radius>=10)
@@ -71,12 +70,12 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
}
}
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,1,1,1),
btVector4(1,1,0,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,1,1,1),
b3Vector4(1,1,0,1),
};
int curColor = 0;
@@ -90,10 +89,10 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(radius,radius,radius,1);
b3Vector4 scaling(radius,radius,radius,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(prevGraphicsShapeIndex,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index, writeInstanceToGpu);
@@ -110,12 +109,12 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
btVector4 colors[4] =
b3Vector4 colors[4] =
{
btVector4(1,0,0,1),
btVector4(0,1,0,1),
btVector4(0,1,1,1),
btVector4(1,1,0,1),
b3Vector4(1,0,0,1),
b3Vector4(0,1,0,1),
b3Vector4(0,1,1,1),
b3Vector4(1,1,0,1),
};
@@ -147,10 +146,10 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
b3Quaternion orn(0,0,0,1);
btVector4 color = colors[curColor];
b3Vector4 color = colors[curColor];
curColor++;
curColor&=3;
btVector4 scaling(radius,radius,radius,1);
b3Vector4 scaling(radius,radius,radius,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(prevGraphicsShapeIndex,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(mass,position,orn,colIndex,index, writeInstanceToGpu);
@@ -163,7 +162,7 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
if (1)
{
int shapeId = ci.m_instancingRenderer->registerShape(&cube_vertices[0],numVertices,cube_indices,numIndices);
btVector4 scaling(0.5,0.5,0.5,1);//1,1,1,1);//0.1,0.1,0.1,1);
b3Vector4 scaling(0.5,0.5,0.5,1);//1,1,1,1);//0.1,0.1,0.1,1);
int colIndex = m_data->m_np->registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
b3Vector3 normal(0,-1,0);
float constant=2;
@@ -175,11 +174,11 @@ void GpuSphereScene::setupScene(const ConstructionInfo& ci)
//int i=0;int j=0;
{
//int colIndex = m_data->m_np->registerPlaneShape(normal,constant);//>registerConvexHullShape(&cube_vertices[0],strideInBytes,numVertices, scaling);
btVector4 position(2*i,70+k*2,2*j+8,0);
b3Vector4 position(2*i,70+k*2,2*j+8,0);
//b3Quaternion orn(0,0,0,1);
b3Quaternion orn(b3Vector3(1,0,0),0.3);
btVector4 color(0,0,1,1);
b3Vector4 color(0,0,1,1);
int id = ci.m_instancingRenderer->registerGraphicsInstance(shapeId,position,orn,color,scaling);
int pid = m_data->m_rigidBodyPipeline->registerPhysicsInstance(1.f,position,orn,colIndex,index,false);

0
demo/gpudemo/rigidbody/GpuSphereScene.h → Demos3/GpuDemos/rigidbody/GpuSphereScene.h
View File

28
demo/gpu_initialize/main.cpp → Demos3/GpuGuiInitialize/main.cpp
View File

@@ -12,7 +12,7 @@
#include "OpenGLWindow/GLInstancingRenderer.h"
//#include "OpenGL3CoreRenderer.h"
#include "Bullet3Common/b3Quickprof.h"
//#include "btGpuDynamicsWorld.h"
//#include "b3GpuDynamicsWorld.h"
#include <assert.h>
#include <string.h>
#include "OpenGLTrueTypeFont/fontstash.h"
@@ -34,7 +34,7 @@
#include "Gwen/Controls/CheckBox.h"
#include "Gwen/Controls/TreeControl.h"
btgDefaultOpenGLWindow* window=0;
b3gDefaultOpenGLWindow* window=0;
GLPrimitiveRenderer* primRenderer = 0;
GwenOpenGL3CoreRenderer* pRenderer = 0;
@@ -159,7 +159,7 @@ public:
for (int i=0;i<numPlatforms;i++)
{
cl_platform_id platform = b3OpenCLUtils::getPlatform(i);
btOpenCLPlatformInfo platformInfo;
b3OpenCLPlatformInfo platformInfo;
b3OpenCLUtils::getPlatformInfo(platform, &platformInfo);
cl_device_type deviceType = CL_DEVICE_TYPE_ALL;
cl_int errNum;
@@ -177,7 +177,7 @@ public:
Gwen::Controls::TreeNode* deviceNode = m_node->AddNode( txt );
cl_device_id device = b3OpenCLUtils::getDevice(context,j);
btOpenCLDeviceInfo info;
b3OpenCLDeviceInfo info;
b3OpenCLUtils::getDeviceInfo(device,&info);
Gwen::Controls::TreeNode* node;
@@ -439,7 +439,7 @@ static void MouseButtonCallback(int button, int state, float x, float y)
sLastmousepos[1] = y;
} else
{
btDefaultMouseButtonCallback(button,state,x,y);
b3DefaultMouseButtonCallback(button,state,x,y);
}
}
@@ -471,9 +471,9 @@ static void MouseMoveCallback(float x,float y)
int main(int argc, char* argv[])
{
window = new btgDefaultOpenGLWindow();
window = new b3gDefaultOpenGLWindow();
btgWindowConstructionInfo wci(sGlutScreenWidth,sGlutScreenHeight);
b3gWindowConstructionInfo wci(sGlutScreenWidth,sGlutScreenHeight);
window->createWindow(wci);
window->setResizeCallback(MyResizeCallback);
@@ -494,8 +494,8 @@ int main(int argc, char* argv[])
do
{
CProfileManager::Reset();
CProfileManager::Increment_Frame_Counter();
b3ProfileManager::Reset();
b3ProfileManager::Increment_Frame_Counter();
window->startRendering();
@@ -511,30 +511,30 @@ int main(int argc, char* argv[])
//saveOpenGLState(sGlutScreenWidth,sGlutScreenHeight);
{
BT_PROFILE("gui->draw");
B3_PROFILE("gui->draw");
//gui->draw(g_OpenGLWidth,g_OpenGLHeight);
{
BT_PROFILE("UpdateText");
B3_PROFILE("UpdateText");
if (prof)
prof->UpdateText();
}
{
BT_PROFILE("RenderCanvas");
B3_PROFILE("RenderCanvas");
pCanvas->RenderCanvas();
}
//restoreOpenGLState();
}
{
BT_PROFILE("window->endRendering");
B3_PROFILE("window->endRendering");
window->endRendering();
}
CProfileManager::dumpAll();
b3ProfileManager::dumpAll();

0
demo/gpu_initialize/premake4.lua → Demos3/GpuGuiInitialize/premake4.lua
View File

0
demo/ObjLoader/list.cpp → Demos3/Wavefront/list.cpp
View File

0
demo/ObjLoader/list.h → Demos3/Wavefront/list.h
View File

0
demo/ObjLoader/objLoader.cpp → Demos3/Wavefront/objLoader.cpp
View File

0
demo/ObjLoader/objLoader.h → Demos3/Wavefront/objLoader.h
View File

0
demo/ObjLoader/objTester.cpp → Demos3/Wavefront/objTester.cpp
View File

0
demo/ObjLoader/obj_parser.cpp → Demos3/Wavefront/obj_parser.cpp
View File

0
demo/ObjLoader/obj_parser.h → Demos3/Wavefront/obj_parser.h
View File

0
demo/ObjLoader/premake4.lua → Demos3/Wavefront/premake4.lua
View File

0
demo/ObjLoader/string_extra.cpp → Demos3/Wavefront/string_extra.cpp
View File

0
demo/ObjLoader/string_extra.h → Demos3/Wavefront/string_extra.h
View File

18
demo/donttouch/Bullet2GpuDemo.cpp → Demos3/donttouch/Bullet2GpuDemo.cpp
View File

@@ -1,8 +1,8 @@
#include "Bullet2GpuDemo.h"
#include "../btGpuDynamicsWorld.h"
#include "../b3GpuDynamicsWorld.h"
#include "GpuRigidBodyDemoInternalData.h"
#include "BulletCollision/CollisionShapes/btBoxShape.h"
#include "gpu_rigidbody/host/btRigidBody.h"
#include "BulletCollision/CollisionShapes/b3BoxShape.h"
#include "gpu_rigidbody/host/b3RigidBody.h"
void Bullet2GpuDemo::setupScene(const ConstructionInfo& ci)
{
@@ -10,14 +10,14 @@ void Bullet2GpuDemo::setupScene(const ConstructionInfo& ci)
// m_data->m_np = np;
// m_data->m_bp = bp;
// m_data->m_rigidBodyPipeline
m_gpuDynamicsWorld = new btGpuDynamicsWorld(m_data->m_bp,m_data->m_np,m_data->m_rigidBodyPipeline);
m_gpuDynamicsWorld = new b3GpuDynamicsWorld(m_data->m_bp,m_data->m_np,m_data->m_rigidBodyPipeline);
btVector3 halfExtents(100,1,100);
btBoxShape* boxShape = new btBoxShape(halfExtents);
btVector3 localInertia;
btScalar mass=1.f;
b3Vector3 halfExtents(100,1,100);
b3BoxShape* boxShape = new b3BoxShape(halfExtents);
b3Vector3 localInertia;
b3Scalar mass=1.f;
boxShape->calculateLocalInertia(mass,localInertia);
btRigidBody* body = new btRigidBody(mass,0,boxShape,localInertia);
b3RigidBody* body = new b3RigidBody(mass,0,boxShape,localInertia);
m_gpuDynamicsWorld->addRigidBody(body);
}

2
demo/donttouch/Bullet2GpuDemo.h → Demos3/donttouch/Bullet2GpuDemo.h
View File

@@ -7,7 +7,7 @@ class Bullet2GpuDemo : public GpuRigidBodyDemo
{
protected:
class btGpuDynamicsWorld* m_gpuDynamicsWorld;
class b3GpuDynamicsWorld* m_gpuDynamicsWorld;
public:

148
demo/donttouch/GpuDemo.cpp → Demos3/donttouch/GpuDemo.cpp
View File

@@ -14,8 +14,8 @@ subject to the following restrictions:
*/
#include "btCpuDynamicsWorld.h"
#include "btGpuDynamicsWorld.h"
#include "b3CpuDynamicsWorld.h"
#include "b3GpuDynamicsWorld.h"
#define SCALING 1.
@@ -27,20 +27,20 @@ subject to the following restrictions:
#include "GpuDemo.h"
//#include "GlutStuff.h"
///btBulletDynamicsCommon.h is the main Bullet include file, contains most common include files.
//#include "btBulletDynamicsCommon.h"
///b3BulletDynamicsCommon.h is the main Bullet include file, contains most common include files.
//#include "b3BulletDynamicsCommon.h"
#include "BulletCollision/CollisionShapes/btTriangleMesh.h"
#include "BulletCollision/CollisionShapes/btBvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/btSphereShape.h"
#include "BulletCollision/CollisionShapes/btConvexHullShape.h"
#include "BulletCollision/CollisionShapes/btBoxShape.h"
#include "BulletCollision/CollisionShapes/btCompoundShape.h"
#include "BulletCollision/CollisionShapes/btStaticPlaneShape.h"
#include "BulletCollision/CollisionShapes/b3TriangleMesh.h"
#include "BulletCollision/CollisionShapes/b3BvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/b3SphereShape.h"
#include "BulletCollision/CollisionShapes/b3ConvexHullShape.h"
#include "BulletCollision/CollisionShapes/b3BoxShape.h"
#include "BulletCollision/CollisionShapes/b3CompoundShape.h"
#include "BulletCollision/CollisionShapes/b3StaticPlaneShape.h"
#include "BulletDynamics/Dynamics/btRigidBody.h"
#include "LinearMath/btDefaultMotionState.h"
#include "BulletDynamics/Dynamics/b3RigidBody.h"
#include "LinearMath/b3DefaultMotionState.h"
#include "LinearMath/b3Quickprof.h"
@@ -61,7 +61,7 @@ void GpuDemo::clientMoveAndDisplay()
if (once)
{
once=false;
btDefaultSerializer* serializer = new btDefaultSerializer();
b3DefaultSerializer* serializer = new b3DefaultSerializer();
m_dynamicsWorld->serialize(serializer);
FILE* file = fopen("testFile.bullet","wb");
@@ -74,7 +74,7 @@ void GpuDemo::clientMoveAndDisplay()
count++;
if (count==25)
{
//CProfileManager::dumpAll();
//b3ProfileManager::dumpAll();
}
}
@@ -103,7 +103,7 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
if (1)
{
btSphereShape* sphere = new btSphereShape(1);
b3SphereShape* sphere = new b3SphereShape(1);
m_collisionShapes.push_back(sphere);
/// Create Dynamic Objects
@@ -129,9 +129,9 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
float gapZ =ci.gapZ;
for(int j = 0;j<sizeZ;j++)
{
//btCollisionShape* shape = k==0? boxShape : colShape;
//b3CollisionShape* shape = k==0? boxShape : colShape;
btCollisionShape* shape = sphere;
b3CollisionShape* shape = sphere;
b3Scalar mass = 1;
@@ -152,9 +152,9 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(startTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(startTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
m_dynamicsWorld->addRigidBody(body);
@@ -167,10 +167,10 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
b3Vector3 planeNormal(0,1,0);
b3Scalar planeConstant=0;
btCollisionShape* shape = new btStaticPlaneShape(planeNormal,planeConstant);
//btBoxShape* plane = new btBoxShape(b3Vector3(100,1,100));
b3CollisionShape* shape = new b3StaticPlaneShape(planeNormal,planeConstant);
//b3BoxShape* plane = new b3BoxShape(b3Vector3(100,1,100));
//plane->initializePolyhedralFeatures();
//btSphereShape* shape = new btSphereShape(1000);
//b3SphereShape* shape = new b3SphereShape(1000);
b3Scalar mass(0.);
@@ -184,9 +184,9 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
groundTransform.setOrigin(b3Vector3(0,0,0));
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(groundTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(groundTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
//add the body to the dynamics world
m_dynamicsWorld->addRigidBody(body);
@@ -197,12 +197,12 @@ void SpheresDemo::setupScene(const ConstructionInfo& ci)
void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
{
btCollisionShape* groundShape =0;
// btCollisionShape* groundShape = new btStaticPlaneShape(b3Vector3(0,1,0),50);
b3CollisionShape* groundShape =0;
// b3CollisionShape* groundShape = new b3StaticPlaneShape(b3Vector3(0,1,0),50);
if (ci.m_useConcaveMesh)
{
btTriangleMesh* meshInterface = new btTriangleMesh();
b3TriangleMesh* meshInterface = new b3TriangleMesh();
b3AlignedObjectArray<b3Vector3> concaveVertices;
concaveVertices.push_back(b3Vector3(0,-20,0));
@@ -217,16 +217,16 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
meshInterface->addTriangle(concaveVertices[0],concaveVertices[4],concaveVertices[1],true);
#if 0
groundShape = new btBvhTriangleMeshShape(meshInterface,true);//btStaticPlaneShape(b3Vector3(0,1,0),50);
groundShape = new b3BvhTriangleMeshShape(meshInterface,true);//b3StaticPlaneShape(b3Vector3(0,1,0),50);
#else
btBoxShape* shape =new btBoxShape(b3Vector3(b3Scalar(250.),b3Scalar(10.),b3Scalar(250.)));
b3BoxShape* shape =new b3BoxShape(b3Vector3(b3Scalar(250.),b3Scalar(10.),b3Scalar(250.)));
shape->initializePolyhedralFeatures();
groundShape = shape;
#endif
} else
{
groundShape = new btBoxShape(b3Vector3(b3Scalar(250.),b3Scalar(50.),b3Scalar(250.)));
groundShape = new b3BoxShape(b3Vector3(b3Scalar(250.),b3Scalar(50.),b3Scalar(250.)));
}
m_collisionShapes.push_back(groundShape);
@@ -248,9 +248,9 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
groundShape->calculateLocalInertia(mass,localInertia);
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(groundTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,groundShape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(groundTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,groundShape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
//add the body to the dynamics world
m_dynamicsWorld->addRigidBody(body);
@@ -272,14 +272,14 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
vertices.push_back(b3Vector3(-1,-1,1));
#if 0
btPolyhedralConvexShape* colShape = new btConvexHullShape(&vertices[0].getX(),vertices.size());
b3PolyhedralConvexShape* colShape = new b3ConvexHullShape(&vertices[0].getX(),vertices.size());
colShape->initializePolyhedralFeatures();
#else
btCompoundShape* compoundShape = 0;
b3CompoundShape* compoundShape = 0;
{
btPolyhedralConvexShape* colShape = new btConvexHullShape(&vertices[0].getX(),vertices.size());
b3PolyhedralConvexShape* colShape = new b3ConvexHullShape(&vertices[0].getX(),vertices.size());
colShape->initializePolyhedralFeatures();
compoundShape = new btCompoundShape();
compoundShape = new b3CompoundShape();
b3Transform tr;
tr.setIdentity();
tr.setOrigin(b3Vector3(0,-1,0));
@@ -289,18 +289,18 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
tr.setOrigin(b3Vector3(2,0,0));
compoundShape->addChildShape(tr,colShape);
}
btCollisionShape* colShape = compoundShape;
b3CollisionShape* colShape = compoundShape;
#endif
btPolyhedralConvexShape* boxShape = new btBoxShape(b3Vector3(SCALING*1,SCALING*1,SCALING*1));
b3PolyhedralConvexShape* boxShape = new b3BoxShape(b3Vector3(SCALING*1,SCALING*1,SCALING*1));
boxShape->initializePolyhedralFeatures();
//btCollisionShape* colShape = new btSphereShape(b3Scalar(1.));
//b3CollisionShape* colShape = new b3SphereShape(b3Scalar(1.));
m_collisionShapes.push_back(colShape);
m_collisionShapes.push_back(boxShape);
@@ -329,9 +329,9 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
float gapZ = !ci.m_useConcaveMesh&&k==0? 3.05 : ci.gapZ;
for(int j = 0;j<sizeZ;j++)
{
//btCollisionShape* shape = k==0? boxShape : colShape;
//b3CollisionShape* shape = k==0? boxShape : colShape;
btCollisionShape* shape = colShape;
b3CollisionShape* shape = colShape;
b3Scalar mass = 1;
@@ -352,9 +352,9 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(startTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(startTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
m_dynamicsWorld->addRigidBody(body);
@@ -368,12 +368,12 @@ void GpuCompoundDemo::setupScene(const ConstructionInfo& ci)
void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
{
btCollisionShape* groundShape =0;
// btCollisionShape* groundShape = new btStaticPlaneShape(b3Vector3(0,1,0),50);
b3CollisionShape* groundShape =0;
// b3CollisionShape* groundShape = new b3StaticPlaneShape(b3Vector3(0,1,0),50);
if (ci.m_useConcaveMesh)
{
btTriangleMesh* meshInterface = new btTriangleMesh();
b3TriangleMesh* meshInterface = new b3TriangleMesh();
b3AlignedObjectArray<b3Vector3> concaveVertices;
concaveVertices.push_back(b3Vector3(0,-20,0));
@@ -388,16 +388,16 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
meshInterface->addTriangle(concaveVertices[0],concaveVertices[4],concaveVertices[1],true);
#if 0
groundShape = new btBvhTriangleMeshShape(meshInterface,true);//btStaticPlaneShape(b3Vector3(0,1,0),50);
groundShape = new b3BvhTriangleMeshShape(meshInterface,true);//b3StaticPlaneShape(b3Vector3(0,1,0),50);
#else
btBoxShape* shape =new btBoxShape(b3Vector3(b3Scalar(250.),b3Scalar(10.),b3Scalar(250.)));
b3BoxShape* shape =new b3BoxShape(b3Vector3(b3Scalar(250.),b3Scalar(10.),b3Scalar(250.)));
shape->initializePolyhedralFeatures();
groundShape = shape;
#endif
} else
{
groundShape = new btBoxShape(b3Vector3(b3Scalar(250.),b3Scalar(50.),b3Scalar(250.)));
groundShape = new b3BoxShape(b3Vector3(b3Scalar(250.),b3Scalar(50.),b3Scalar(250.)));
}
m_collisionShapes.push_back(groundShape);
@@ -419,9 +419,9 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
groundShape->calculateLocalInertia(mass,localInertia);
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(groundTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,groundShape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(groundTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,groundShape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
//add the body to the dynamics world
m_dynamicsWorld->addRigidBody(body);
@@ -443,14 +443,14 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
vertices.push_back(b3Vector3(-1,-1,1));
#if 1
btPolyhedralConvexShape* colShape = new btConvexHullShape(&vertices[0].getX(),vertices.size());
b3PolyhedralConvexShape* colShape = new b3ConvexHullShape(&vertices[0].getX(),vertices.size());
colShape->initializePolyhedralFeatures();
#else
btCompoundShape* compoundShape = 0;
b3CompoundShape* compoundShape = 0;
{
btPolyhedralConvexShape* colShape = new btConvexHullShape(&vertices[0].getX(),vertices.size());
b3PolyhedralConvexShape* colShape = new b3ConvexHullShape(&vertices[0].getX(),vertices.size());
colShape->initializePolyhedralFeatures();
compoundShape = new btCompoundShape();
compoundShape = new b3CompoundShape();
b3Transform tr;
tr.setIdentity();
tr.setOrigin(b3Vector3(0,-1,0));
@@ -460,18 +460,18 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
tr.setOrigin(b3Vector3(2,0,0));
compoundShape->addChildShape(tr,colShape);
}
btCollisionShape* colShape = compoundShape;
b3CollisionShape* colShape = compoundShape;
#endif
btPolyhedralConvexShape* boxShape = new btBoxShape(b3Vector3(SCALING*1,SCALING*1,SCALING*1));
b3PolyhedralConvexShape* boxShape = new b3BoxShape(b3Vector3(SCALING*1,SCALING*1,SCALING*1));
boxShape->initializePolyhedralFeatures();
//btCollisionShape* colShape = new btSphereShape(b3Scalar(1.));
//b3CollisionShape* colShape = new b3SphereShape(b3Scalar(1.));
m_collisionShapes.push_back(colShape);
m_collisionShapes.push_back(boxShape);
@@ -500,9 +500,9 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
float gapZ = !ci.m_useConcaveMesh&&k==0? 3.05 : ci.gapZ;
for(int j = 0;j<sizeZ;j++)
{
//btCollisionShape* shape = k==0? boxShape : colShape;
//b3CollisionShape* shape = k==0? boxShape : colShape;
btCollisionShape* shape = colShape;
b3CollisionShape* shape = colShape;
b3Scalar mass = 1;
@@ -523,9 +523,9 @@ void GpuBoxDemo::setupScene(const ConstructionInfo& ci)
//using motionstate is recommended, it provides interpolation capabilities, and only synchronizes 'active' objects
btDefaultMotionState* myMotionState = new btDefaultMotionState(startTransform);
btRigidBody::btRigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
btRigidBody* body = new btRigidBody(rbInfo);
b3DefaultMotionState* myMotionState = new b3DefaultMotionState(startTransform);
b3RigidBody::b3RigidBodyConstructionInfo rbInfo(mass,myMotionState,shape,localInertia);
b3RigidBody* body = new b3RigidBody(rbInfo);
m_dynamicsWorld->addRigidBody(body);
@@ -546,10 +546,10 @@ void GpuDemo::initPhysics(const ConstructionInfo& ci)
///collision configuration contains default setup for memory, collision setup
if (ci.useOpenCL)
{
m_dynamicsWorld = new btGpuDynamicsWorld(ci.preferredOpenCLPlatformIndex,ci.preferredOpenCLDeviceIndex);
m_dynamicsWorld = new b3GpuDynamicsWorld(ci.preferredOpenCLPlatformIndex,ci.preferredOpenCLDeviceIndex);
} else
{
m_dynamicsWorld = new btCpuDynamicsWorld();
m_dynamicsWorld = new b3CpuDynamicsWorld();
}
@@ -581,8 +581,8 @@ void GpuDemo::exitPhysics()
{
for (i=m_dynamicsWorld->getNumCollisionObjects()-1; i>=0 ;i--)
{
btCollisionObject* obj = m_dynamicsWorld->getCollisionObjectArray()[i];
btRigidBody* body = btRigidBody::upcast(obj);
b3CollisionObject* obj = m_dynamicsWorld->getCollisionObjectArray()[i];
b3RigidBody* body = b3RigidBody::upcast(obj);
if (body && body->getMotionState())
{
delete body->getMotionState();
@@ -595,7 +595,7 @@ void GpuDemo::exitPhysics()
//delete collision shapes
for (int j=0;j<m_collisionShapes.size();j++)
{
btCollisionShape* shape = m_collisionShapes[j];
b3CollisionShape* shape = m_collisionShapes[j];
delete shape;
}
m_collisionShapes.clear();

20
demo/donttouch/GpuDemo.h → Demos3/donttouch/GpuDemo.h
View File

@@ -19,14 +19,14 @@ subject to the following restrictions:
#include "Bullet3Common/b3AlignedObjectArray.h"
class btBroadphaseInterface;
class btCollisionShape;
class b3BroadphaseInterface;
class b3CollisionShape;
class b3OverlappingPairCache;
class btCollisionDispatcher;
class btConstraintSolver;
struct btCollisionAlgorithmCreateFunc;
class btDefaultCollisionConfiguration;
class btDynamicsWorld;
class b3CollisionDispatcher;
class b3ConstraintSolver;
struct b3CollisionAlgorithmCreateFunc;
class b3DefaultCollisionConfiguration;
class b3DynamicsWorld;
class GLInstancingRenderer;
///GpuDemo is good starting point for learning the code base and porting.
@@ -38,10 +38,10 @@ class GpuDemo
protected:
btDynamicsWorld* m_dynamicsWorld;
b3DynamicsWorld* m_dynamicsWorld;
//keep the collision shapes, for deletion/cleanup
b3AlignedObjectArray<btCollisionShape*> m_collisionShapes;
b3AlignedObjectArray<b3CollisionShape*> m_collisionShapes;
float getDeltaTimeInSeconds()
{
@@ -99,7 +99,7 @@ public:
virtual void exitPhysics();
virtual const btDynamicsWorld* getDynamicsWorld() const
virtual const b3DynamicsWorld* getDynamicsWorld() const
{
return m_dynamicsWorld;
}

70
demo/donttouch/OpenGL3CoreRenderer.cpp → Demos3/donttouch/OpenGL3CoreRenderer.cpp
View File

@@ -2,18 +2,18 @@
#include "OpenGL3CoreRenderer.h"
#include "OpenGLWindow/GLInstancingRenderer.h"
#include "OpenGLWindow/ShapeData.h"
//#include "BulletDynamics/Dynamics/btDiscreteDynamicsWorld.h"
//#include "BulletCollision/CollisionDispatch/btCollisionObject.h"
//#include "BulletDynamics/Dynamics/b3DiscreteDynamicsWorld.h"
//#include "BulletCollision/CollisionDispatch/b3CollisionObject.h"
#include "Bullet3Common/b3Quickprof.h"
/*#include "BulletCollision/CollisionShapes/btBvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/btConvexPolyhedron.h"
#include "BulletCollision/CollisionShapes/btConvexHullShape.h"
#include "BulletCollision/CollisionShapes/btCollisionShape.h"
#include "BulletCollision/CollisionShapes/btBoxShape.h"
#include "BulletCollision/CollisionShapes/btCompoundShape.h"
#include "BulletCollision/CollisionShapes/btSphereShape.h"
#include "BulletCollision/CollisionShapes/btStaticPlaneShape.h"
/*#include "BulletCollision/CollisionShapes/b3BvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/b3ConvexPolyhedron.h"
#include "BulletCollision/CollisionShapes/b3ConvexHullShape.h"
#include "BulletCollision/CollisionShapes/b3CollisionShape.h"
#include "BulletCollision/CollisionShapes/b3BoxShape.h"
#include "BulletCollision/CollisionShapes/b3CompoundShape.h"
#include "BulletCollision/CollisionShapes/b3SphereShape.h"
#include "BulletCollision/CollisionShapes/b3StaticPlaneShape.h"
#include "../../rendering/WavefrontObjLoader/objLoader.h"
*/
@@ -61,7 +61,7 @@ struct GraphicsShape
GraphicsShape* createGraphicsShapeFromConvexHull(const btConvexPolyhedron* utilPtr)
GraphicsShape* createGraphicsShapeFromConvexHull(const b3ConvexPolyhedron* utilPtr)
{
b3AlignedObjectArray<GraphicsVertex>* vertices = new b3AlignedObjectArray<GraphicsVertex>;
@@ -71,7 +71,7 @@ GraphicsShape* createGraphicsShapeFromConvexHull(const btConvexPolyhedron* utilP
b3AlignedObjectArray<int>* indicesPtr = new b3AlignedObjectArray<int>;
for (int f=0;f<utilPtr->m_faces.size();f++)
{
const btFace& face = utilPtr->m_faces[f];
const b3Face& face = utilPtr->m_faces[f];
b3Vector3 normal(face.m_plane[0],face.m_plane[1],face.m_plane[2]);
if (face.m_indices.size()>2)
{
@@ -124,7 +124,7 @@ GraphicsShape* createGraphicsShapeFromConvexHull(const btConvexPolyhedron* utilP
}
}
GraphicsShape* createGraphicsShapeFromCompoundShape(btCompoundShape* compound)
GraphicsShape* createGraphicsShapeFromCompoundShape(b3CompoundShape* compound)
{
GraphicsShape* gfxShape = new GraphicsShape();
b3AlignedObjectArray<GraphicsVertex>* vertexArray = new b3AlignedObjectArray<GraphicsVertex>;
@@ -135,13 +135,13 @@ GraphicsShape* createGraphicsShapeFromCompoundShape(btCompoundShape* compound)
//create a graphics shape for each child, combine them into a single graphics shape using their child transforms
for (int i=0;i<compound->getNumChildShapes();i++)
{
btAssert(compound->getChildShape(i)->isPolyhedral());
b3Assert(compound->getChildShape(i)->isPolyhedral());
if (compound->getChildShape(i)->isPolyhedral())
{
btPolyhedralConvexShape* convexHull = (btPolyhedralConvexShape*) compound->getChildShape(i);
b3PolyhedralConvexShape* convexHull = (b3PolyhedralConvexShape*) compound->getChildShape(i);
b3Transform tr = compound->getChildTransform(i);
const btConvexPolyhedron* polyhedron = convexHull->getConvexPolyhedron();
const b3ConvexPolyhedron* polyhedron = convexHull->getConvexPolyhedron();
GraphicsShape* childGfxShape = createGraphicsShapeFromConvexHull(polyhedron);
int baseIndex = vertexArray->size();
@@ -173,8 +173,8 @@ GraphicsShape* createGraphicsShapeFromCompoundShape(btCompoundShape* compound)
}
}
btPolyhedralConvexShape* convexHull = (btPolyhedralConvexShape*) compound->getChildShape(0);
const btConvexPolyhedron* polyhedron = convexHull->getConvexPolyhedron();
b3PolyhedralConvexShape* convexHull = (b3PolyhedralConvexShape*) compound->getChildShape(0);
const b3ConvexPolyhedron* polyhedron = convexHull->getConvexPolyhedron();
GraphicsShape* childGfxShape = createGraphicsShapeFromConvexHull(polyhedron);
gfxShape->m_indices = &indexArray->at(0);
@@ -189,7 +189,7 @@ GraphicsShape* createGraphicsShapeFromCompoundShape(btCompoundShape* compound)
return gfxShape;
}
GraphicsShape* createGraphicsShapeFromConcaveMesh(const btBvhTriangleMeshShape* trimesh)
GraphicsShape* createGraphicsShapeFromConcaveMesh(const b3BvhTriangleMeshShape* trimesh)
{
b3AlignedObjectArray<GraphicsVertex>* vertices = new b3AlignedObjectArray<GraphicsVertex>;
@@ -311,7 +311,7 @@ GraphicsShape* createGraphicsShapeFromWavefrontObj(objLoader* obj)
vtx.xyzw[1] = obj->vertexList[v]->e[1];
vtx.xyzw[2] = obj->vertexList[v]->e[2];
b3Vector3 n(vtx.xyzw[0],vtx.xyzw[1],vtx.xyzw[2]);
if (n.length2()>SIMD_EPSILON)
if (n.length2()>B3_EPSILON)
{
n.normalize();
vtx.normal[0] = n[0];
@@ -425,15 +425,15 @@ GraphicsShape* createGraphicsShapeFromWavefrontObj(objLoader* obj)
//very incomplete conversion from physics to graphics
void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOnly, int numObjects, btCollisionObject** colObjArray)
void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOnly, int numObjects, b3CollisionObject** colObjArray)
{
///@todo: we need to sort the objects based on collision shape type, so we can share instances
BT_PROFILE("graphics_from_physics");
B3_PROFILE("graphics_from_physics");
int strideInBytes = sizeof(float)*9;
int prevGraphicsShapeIndex = -1;
btCollisionShape* prevShape = 0;
b3CollisionShape* prevShape = 0;
int numColObj = numObjects;
@@ -444,7 +444,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
for (int i=0;i<numColObj;i++)
{
btCollisionObject* colObj = colObjArray[i];
b3CollisionObject* colObj = colObjArray[i];
b3Vector3 pos = colObj->getWorldTransform().getOrigin();
b3Quaternion orn = colObj->getWorldTransform().getRotation();
@@ -470,8 +470,8 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
if (colObj->getCollisionShape()->isPolyhedral())
{
btPolyhedralConvexShape* polyShape = (btPolyhedralConvexShape*)colObj->getCollisionShape();
const btConvexPolyhedron* pol = polyShape->getConvexPolyhedron();
b3PolyhedralConvexShape* polyShape = (b3PolyhedralConvexShape*)colObj->getCollisionShape();
const b3ConvexPolyhedron* pol = polyShape->getConvexPolyhedron();
GraphicsShape* gfxShape = createGraphicsShapeFromConvexHull(pol);
prevGraphicsShapeIndex = renderer.registerShape(&gfxShape->m_vertices[0],gfxShape->m_numvertices,gfxShape->m_indices,gfxShape->m_numIndices);
@@ -482,7 +482,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
if (colObj->getCollisionShape()->getShapeType()==TRIANGLE_MESH_SHAPE_PROXYTYPE)
{
btBvhTriangleMeshShape* trimesh = (btBvhTriangleMeshShape*) colObj->getCollisionShape();
b3BvhTriangleMeshShape* trimesh = (b3BvhTriangleMeshShape*) colObj->getCollisionShape();
GraphicsShape* gfxShape = createGraphicsShapeFromConcaveMesh(trimesh);
prevGraphicsShapeIndex = renderer.registerShape(&gfxShape->m_vertices[0],gfxShape->m_numvertices,gfxShape->m_indices,gfxShape->m_numIndices);
prevShape = colObj->getCollisionShape();
@@ -492,7 +492,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
if (colObj->getCollisionShape()->getShapeType()==COMPOUND_SHAPE_PROXYTYPE)
{
btCompoundShape* compound = (btCompoundShape*) colObj->getCollisionShape();
b3CompoundShape* compound = (b3CompoundShape*) colObj->getCollisionShape();
GraphicsShape* gfxShape = createGraphicsShapeFromCompoundShape(compound);
if (gfxShape)
{
@@ -508,11 +508,11 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
if (colObj->getCollisionShape()->getShapeType()==SPHERE_SHAPE_PROXYTYPE)
{
btSphereShape* sphere = (btSphereShape*) colObj->getCollisionShape();
b3SphereShape* sphere = (b3SphereShape*) colObj->getCollisionShape();
b3Scalar radius = sphere->getRadius();
//btConvexHullShape* spherePoly = new btConvexHullShape(
//const btConvexPolyhedron* pol = polyShape->getConvexPolyhedron();
//b3ConvexHullShape* spherePoly = new b3ConvexHullShape(
//const b3ConvexPolyhedron* pol = polyShape->getConvexPolyhedron();
/*
objLoader loader;
@@ -572,7 +572,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
int numVertices = sizeof(point_sphere_vertices)/strideInBytes;
int numIndices = sizeof(point_sphere_indices)/sizeof(int);
prevGraphicsShapeIndex = renderer.registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,BT_GL_POINTS);
prevGraphicsShapeIndex = renderer.registerShape(&point_sphere_vertices[0],numVertices,point_sphere_indices,numIndices,B3_GL_POINTS);
} else
{
if (radius>=10)
@@ -599,7 +599,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
if (colObj->getCollisionShape()->getShapeType()==STATIC_PLANE_PROXYTYPE)
{
btStaticPlaneShape* plane= (btStaticPlaneShape*) colObj->getCollisionShape();
b3StaticPlaneShape* plane= (b3StaticPlaneShape*) colObj->getCollisionShape();
prevShape = colObj->getCollisionShape();
//plane->getPlaneNormal()
@@ -624,7 +624,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
{
printf("Error: unsupported collision shape type in %s %d\n", __FILE__, __LINE__);
prevGraphicsShapeIndex = -1;
btAssert(0);
b3Assert(0);
}
}
}
@@ -660,7 +660,7 @@ void graphics_from_physics(GLInstancingRenderer& renderer, bool syncTransformsOn
void OpenGL3CoreRenderer::renderPhysicsWorld(int numObjects, btCollisionObject** colObjArray, bool syncOnly)
void OpenGL3CoreRenderer::renderPhysicsWorld(int numObjects, b3CollisionObject** colObjArray, bool syncOnly)
{
//sync changes from physics world to render world
//for now, we don't deal with adding/removing objects to the world during the simulation, to keep the rendererer simpler

4
demo/donttouch/OpenGL3CoreRenderer.h → Demos3/donttouch/OpenGL3CoreRenderer.h
View File

@@ -1,7 +1,7 @@
#ifndef OPENGL3_CORE_RENDERER_H
#define OPENGL3_CORE_RENDERER_H
class btCollisionObject;
class b3CollisionObject;
class GLInstancingRenderer;
class OpenGL3CoreRenderer
@@ -14,7 +14,7 @@ public:
void init();
void reshape(int w, int h);
void keyboardCallback(unsigned char key);
void renderPhysicsWorld(int numObjects, btCollisionObject** colObjArray, bool syncOnly);
void renderPhysicsWorld(int numObjects, b3CollisionObject** colObjArray, bool syncOnly);
GLInstancingRenderer* getInstancingRenderer()
{

17
Demos3/donttouch/b3CpuDynamicsWorld.cpp Normal file
View File

@@ -0,0 +1,17 @@
#include "b3CpuDynamicsWorld.h"
#include "b3BulletDynamicsCommon.h"
b3CpuDynamicsWorld::b3CpuDynamicsWorld()
:b3DiscreteDynamicsWorld(
new b3CollisionDispatcher(new b3DefaultCollisionConfiguration()),
new b3DynamicBvhBroadphase(),new b3SequentialImpulseConstraintSolver(),
new b3DefaultCollisionConfiguration()//todo: remove this!
)
{
}
b3CpuDynamicsWorld::~b3CpuDynamicsWorld()
{
}

24
Demos3/donttouch/b3CpuDynamicsWorld.h Normal file
View File

@@ -0,0 +1,24 @@
#ifndef B3_CPU_DYNAMICS_WORLD_H
#define B3_CPU_DYNAMICS_WORLD_H
class b3DefaultCollisionConfiguration;
class b3CollisionDispatcher;
struct b3DynamicBvhBroadphase;
class b3SequentialImpulseConstraintSolver;
#include "BulletDynamics/Dynamics/b3DiscreteDynamicsWorld.h"
class b3CpuDynamicsWorld : public b3DiscreteDynamicsWorld
{
public:
b3CpuDynamicsWorld();
virtual ~b3CpuDynamicsWorld();
};
#endif //B3_CPU_DYNAMICS_WORLD_H

86
demo/donttouch/btGpuDynamicsWorld.cpp → Demos3/donttouch/b3GpuDynamicsWorld.cpp
View File

@@ -1,13 +1,13 @@
#include "btGpuDynamicsWorld.h"
#include "BulletDynamics/Dynamics/btRigidBody.h"
#include "b3GpuDynamicsWorld.h"
#include "BulletDynamics/Dynamics/b3RigidBody.h"
#include "../../../opencl/gpu_rigidbody_pipeline2/CLPhysicsDemo.h"
#include "../../../opencl/gpu_rigidbody_pipeline/btGpuNarrowPhaseAndSolver.h"
#include "BulletCollision/CollisionShapes/btPolyhedralConvexShape.h"
#include "BulletCollision/CollisionShapes/btBvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/btCompoundShape.h"
#include "BulletCollision/CollisionShapes/btSphereShape.h"
#include "BulletCollision/CollisionShapes/btStaticPlaneShape.h"
#include "../../../opencl/gpu_rigidbody_pipeline/b3GpuNarrowPhaseAndSolver.h"
#include "BulletCollision/CollisionShapes/b3PolyhedralConvexShape.h"
#include "BulletCollision/CollisionShapes/b3BvhTriangleMeshShape.h"
#include "BulletCollision/CollisionShapes/b3CompoundShape.h"
#include "BulletCollision/CollisionShapes/b3SphereShape.h"
#include "BulletCollision/CollisionShapes/b3StaticPlaneShape.h"
#include "LinearMath/b3Quickprof.h"
@@ -18,8 +18,8 @@
btGpuDynamicsWorld::btGpuDynamicsWorld(int preferredOpenCLPlatformIndex,int preferredOpenCLDeviceIndex)
:btDynamicsWorld(0,0,0),
b3GpuDynamicsWorld::b3GpuDynamicsWorld(int preferredOpenCLPlatformIndex,int preferredOpenCLDeviceIndex)
:b3DynamicsWorld(0,0,0),
m_gravity(0,-10,0),
m_once(true)
{
@@ -29,12 +29,12 @@ m_once(true)
m_gpuPhysics->init(preferredOpenCLDeviceIndex,preferredOpenCLPlatformIndex,useInterop);
}
btGpuDynamicsWorld::~btGpuDynamicsWorld()
b3GpuDynamicsWorld::~b3GpuDynamicsWorld()
{
delete m_gpuPhysics;
}
void btGpuDynamicsWorld::exitOpenCL()
void b3GpuDynamicsWorld::exitOpenCL()
{
}
@@ -43,13 +43,13 @@ void btGpuDynamicsWorld::exitOpenCL()
int btGpuDynamicsWorld::stepSimulation( b3Scalar timeStep,int maxSubSteps, b3Scalar fixedTimeStep)
int b3GpuDynamicsWorld::stepSimulation( b3Scalar timeStep,int maxSubSteps, b3Scalar fixedTimeStep)
{
#ifndef BT_NO_PROFILE
// CProfileManager::Reset();
#endif //BT_NO_PROFILE
#ifndef B3_NO_PROFILE
// b3ProfileManager::Reset();
#endif //B3_NO_PROFILE
BT_PROFILE("stepSimulation");
B3_PROFILE("stepSimulation");
//convert all shapes now, and if any change, reset all (todo)
@@ -63,14 +63,14 @@ int btGpuDynamicsWorld::stepSimulation( b3Scalar timeStep,int maxSubSteps, b3Sc
{
{
BT_PROFILE("readbackBodiesToCpu");
B3_PROFILE("readbackBodiesToCpu");
//now copy info back to rigid bodies....
m_gpuPhysics->readbackBodiesToCpu();
}
{
BT_PROFILE("scatter transforms into rigidbody (CPU)");
B3_PROFILE("scatter transforms into rigidbody (CPU)");
for (int i=0;i<this->m_collisionObjects.size();i++)
{
b3Vector3 pos;
@@ -85,20 +85,20 @@ int btGpuDynamicsWorld::stepSimulation( b3Scalar timeStep,int maxSubSteps, b3Sc
}
#ifndef BT_NO_PROFILE
//CProfileManager::Increment_Frame_Counter();
#endif //BT_NO_PROFILE
#ifndef B3_NO_PROFILE
//b3ProfileManager::Increment_Frame_Counter();
#endif //B3_NO_PROFILE
return 1;
}
void btGpuDynamicsWorld::setGravity(const b3Vector3& gravity)
void b3GpuDynamicsWorld::setGravity(const b3Vector3& gravity)
{
}
int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* colShape)
int b3GpuDynamicsWorld::findOrRegisterCollisionShape(const b3CollisionShape* colShape)
{
int index = m_uniqueShapes.findLinearSearch(colShape);
if (index==m_uniqueShapes.size())
@@ -107,7 +107,7 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
{
m_uniqueShapes.push_back(colShape);
btPolyhedralConvexShape* convex = (btPolyhedralConvexShape*)colShape;
b3PolyhedralConvexShape* convex = (b3PolyhedralConvexShape*)colShape;
int numVertices=convex->getNumVertices();
int strideInBytes=sizeof(b3Vector3);
@@ -126,7 +126,7 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
{
m_uniqueShapes.push_back(colShape);
btBvhTriangleMeshShape* trimesh = (btBvhTriangleMeshShape*) colShape;
b3BvhTriangleMeshShape* trimesh = (b3BvhTriangleMeshShape*) colShape;
b3StridingMeshInterface* meshInterface = trimesh->getMeshInterface();
b3AlignedObjectArray<b3Vector3> vertices;
b3AlignedObjectArray<int> indices;
@@ -181,9 +181,9 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
indices.push_back(indices.size());
}
}
//GraphicsShape* gfxShape = 0;//btBulletDataExtractor::createGraphicsShapeFromWavefrontObj(objData);
//GraphicsShape* gfxShape = 0;//b3BulletDataExtractor::createGraphicsShapeFromWavefrontObj(objData);
//GraphicsShape* gfxShape = btBulletDataExtractor::createGraphicsShapeFromConvexHull(&sUnitSpherePoints[0],MY_UNITSPHERE_POINTS);
//GraphicsShape* gfxShape = b3BulletDataExtractor::createGraphicsShapeFromConvexHull(&sUnitSpherePoints[0],MY_UNITSPHERE_POINTS);
float meshScaling[4] = {1,1,1,1};
//int shapeIndex = renderer.registerShape(gfxShape->m_vertices,gfxShape->m_numvertices,gfxShape->m_indices,gfxShape->m_numIndices);
float groundPos[4] = {0,0,0,0};
@@ -195,9 +195,9 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
m_uniqueShapeMapping.push_back(gpuShapeIndex);
} else
{
printf("Error: no vertices in mesh in btGpuDynamicsWorld::addRigidBody\n");
printf("Error: no vertices in mesh in b3GpuDynamicsWorld::addRigidBody\n");
index = -1;
btAssert(0);
b3Assert(0);
}
@@ -206,14 +206,14 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
if (colShape->getShapeType()==COMPOUND_SHAPE_PROXYTYPE)
{
btCompoundShape* compound = (btCompoundShape*) colShape;
b3AlignedObjectArray<btGpuChildShape> childShapes;
b3CompoundShape* compound = (b3CompoundShape*) colShape;
b3AlignedObjectArray<b3GpuChildShape> childShapes;
for (int i=0;i<compound->getNumChildShapes();i++)
{
//for now, only support polyhedral child shapes
btAssert(compound->getChildShape(i)->isPolyhedral());
btGpuChildShape child;
b3Assert(compound->getChildShape(i)->isPolyhedral());
b3GpuChildShape child;
child.m_shapeIndex = findOrRegisterCollisionShape(compound->getChildShape(i));
b3Vector3 pos = compound->getChildTransform(i).getOrigin();
b3Quaternion orn = compound->getChildTransform(i).getRotation();
@@ -233,16 +233,16 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
/*printf("Error: unsupported compound type (%d) in btGpuDynamicsWorld::addRigidBody\n",colShape->getShapeType());
/*printf("Error: unsupported compound type (%d) in b3GpuDynamicsWorld::addRigidBody\n",colShape->getShapeType());
index = -1;
btAssert(0);
b3Assert(0);
*/
} else
{
if (colShape->getShapeType()==SPHERE_SHAPE_PROXYTYPE)
{
m_uniqueShapes.push_back(colShape);
btSphereShape* sphere = (btSphereShape*)colShape;
b3SphereShape* sphere = (b3SphereShape*)colShape;
int gpuShapeIndex = m_gpuPhysics->registerSphereShape(sphere->getRadius());
m_uniqueShapeMapping.push_back(gpuShapeIndex);
@@ -251,15 +251,15 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
if (colShape->getShapeType()==STATIC_PLANE_PROXYTYPE)
{
m_uniqueShapes.push_back(colShape);
btStaticPlaneShape* plane = (btStaticPlaneShape*)colShape;
b3StaticPlaneShape* plane = (b3StaticPlaneShape*)colShape;
int gpuShapeIndex = m_gpuPhysics->registerPlaneShape(plane->getPlaneNormal(),plane->getPlaneConstant());
m_uniqueShapeMapping.push_back(gpuShapeIndex);
} else
{
printf("Error: unsupported shape type (%d) in btGpuDynamicsWorld::addRigidBody\n",colShape->getShapeType());
printf("Error: unsupported shape type (%d) in b3GpuDynamicsWorld::addRigidBody\n",colShape->getShapeType());
index = -1;
btAssert(0);
b3Assert(0);
}
}
}
@@ -271,7 +271,7 @@ int btGpuDynamicsWorld::findOrRegisterCollisionShape(const btCollisionShape* col
return index;
}
void btGpuDynamicsWorld::addRigidBody(btRigidBody* body)
void b3GpuDynamicsWorld::addRigidBody(b3RigidBody* body)
{
body->setMotionState(0);
@@ -292,9 +292,9 @@ void btGpuDynamicsWorld::addRigidBody(btRigidBody* body)
}
}
void btGpuDynamicsWorld::removeCollisionObject(btCollisionObject* colObj)
void b3GpuDynamicsWorld::removeCollisionObject(b3CollisionObject* colObj)
{
btDynamicsWorld::removeCollisionObject(colObj);
b3DynamicsWorld::removeCollisionObject(colObj);
}

109
Demos3/donttouch/b3GpuDynamicsWorld.h Normal file
View File

@@ -0,0 +1,109 @@
#ifndef B3_GPU_DYNAMICS_WORLD_H
#define B3_GPU_DYNAMICS_WORLD_H
class b3Vector3;
class b3RigidBody;
class b3CollisionObject;
struct b3GpuInternalData;//use this struct to avoid 'leaking' all OpenCL headers into clients code base
class CLPhysicsDemo;
#include "Bullet3Common/b3AlignedObjectArray.h"
//#include "BulletDynamics/Dynamics/b3DynamicsWorld.h"
class b3GpuDynamicsWorld //: public b3DynamicsWorld
{
b3AlignedObjectArray<const class b3CollisionShape*> m_uniqueShapes;
b3AlignedObjectArray<int> m_uniqueShapeMapping;
CLPhysicsDemo* m_gpuPhysics;
b3Vector3 m_gravity;
bool m_once;
bool initOpenCL(int preferredDeviceIndex, int preferredPlatformIndex, bool useInterop);
void exitOpenCL();
int findOrRegisterCollisionShape(const b3CollisionShape* colShape);
public:
b3GpuDynamicsWorld(int preferredOpenCLPlatformIndex,int preferredOpenCLDeviceIndex);
virtual ~b3GpuDynamicsWorld();
virtual int stepSimulation( b3Scalar timeStep,int maxSubSteps=1, b3Scalar fixedTimeStep=b3Scalar(1.)/b3Scalar(60.));
virtual void synchronizeMotionStates()
{
b3Assert(0);
}
void debugDrawWorld() {}
void setGravity(const b3Vector3& gravity);
void addRigidBody(b3RigidBody* body);
void removeCollisionObject(b3CollisionObject* colObj);
b3AlignedObjectArray<class b3CollisionObject*>& getCollisionObjectArray();
const b3AlignedObjectArray<class b3CollisionObject*>& getCollisionObjectArray() const;
virtual void addAction(b3ActionInterface* action)
{
b3Assert(0);
}
virtual void removeAction(b3ActionInterface* action)
{
b3Assert(0);
}
b3Vector3 getGravity () const
{
return m_gravity;
}
virtual void addRigidBody(b3RigidBody* body, short group, short mask)
{
addRigidBody(body);
}
virtual void removeRigidBody(b3RigidBody* body)
{
b3Assert(0);
}
virtual void setConstraintSolver(b3ConstraintSolver* solver)
{
b3Assert(0);
}
virtual b3ConstraintSolver* getConstraintSolver()
{
b3Assert(0);
return 0;
}
virtual b3DynamicsWorldType getWorldType() const
{
return B3_GPU_PHYSICS_WORLD;
}
virtual void clearForces()
{
b3Assert(0);
}
};
#endif //B3_GPU_DYNAMICS_WORLD_H

8
btgui/GwenOpenGLTest/OpenGLSample.cpp
View File

@@ -30,7 +30,7 @@ Gwen::Controls::Canvas* pCanvas = NULL;
void MyMouseMoveCallback( float x, float y)
{
//btDefaultMouseCallback(button,state,x,y);
//b3DefaultMouseCallback(button,state,x,y);
static int m_lastmousepos[2] = {0,0};
static bool isInitialized = false;
@@ -48,7 +48,7 @@ void MyMouseMoveCallback( float x, float y)
void MyMouseButtonCallback(int button, int state, float x, float y)
{
//btDefaultMouseCallback(button,state,x,y);
//b3DefaultMouseCallback(button,state,x,y);
if (pCanvas)
{
@@ -219,8 +219,8 @@ int main()
float retinaScale = 1.f;
btgDefaultOpenGLWindow* window = new btgDefaultOpenGLWindow();
btgWindowConstructionInfo wci;
b3gDefaultOpenGLWindow* window = new b3gDefaultOpenGLWindow();
b3gWindowConstructionInfo wci;
wci.m_width = sWidth;
wci.m_height = sHeight;

2
btgui/GwenOpenGLTest/premake4.lua
View File

@@ -1,5 +1,5 @@
project "Gwen_OpenGLTest"
project "Test_Gwen_OpenGL"
kind "ConsoleApp"
flags {"Unicode"}

134
btgui/OpenGLTrueTypeFont/main.cpp
View File

@@ -94,7 +94,7 @@ void loadBufferData(){
glBindBuffer(GL_ARRAY_BUFFER, sData.m_vertexBuffer);
glBufferData(GL_ARRAY_BUFFER, 4 * sizeof(Vertex), vertexData, GL_STATIC_DRAW);
GLuint err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
@@ -105,7 +105,7 @@ void loadBufferData(){
glEnableVertexAttribArray(sData.m_positionAttribute);
glEnableVertexAttribArray(sData.m_colourAttribute);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glEnableVertexAttribArray(sData.m_textureAttribute);
@@ -113,7 +113,7 @@ void loadBufferData(){
glVertexAttribPointer(sData.m_colourAttribute , 4, GL_FLOAT, GL_FALSE, sizeof(Vertex), (const GLvoid *)sizeof(vec4));
glVertexAttribPointer(sData.m_textureAttribute , 2, GL_FLOAT, GL_FALSE, sizeof(Vertex), (const GLvoid *)(sizeof(vec4)+sizeof(vec4)));
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
}
@@ -123,15 +123,15 @@ void initTestTexture()
glGenTextures(1,(GLuint*)&sData.m_texturehandle);
GLint err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glBindTexture(GL_TEXTURE_2D,sData.m_texturehandle);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
int width=256;
int height=256;
@@ -150,12 +150,12 @@ void initTestTexture()
glTexImage2D(GL_TEXTURE_2D, 0, GL_RED, width,height,0,GL_RED,GL_UNSIGNED_BYTE,image);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glGenerateMipmap(GL_TEXTURE_2D);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
free(image);
@@ -206,19 +206,19 @@ void loadShader(){
sData.m_positionUniform = glGetUniformLocation(sData.m_shaderProg, "p");
if (sData.m_positionUniform < 0) {
btAssert(0);
b3Assert(0);
}
sData.m_colourAttribute = glGetAttribLocation(sData.m_shaderProg, "colour");
if (sData.m_colourAttribute < 0) {
btAssert(0);
b3Assert(0);
}
sData.m_positionAttribute = glGetAttribLocation(sData.m_shaderProg, "position");
if (sData.m_positionAttribute < 0) {
btAssert(0);
b3Assert(0);
}
sData.m_textureAttribute = glGetAttribLocation(sData.m_shaderProg,"texuv");
if (sData.m_textureAttribute < 0) {
btAssert(0);
b3Assert(0);
}
}
@@ -227,7 +227,7 @@ void display() {
GLint err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
const float timeScale = 0.008f;
@@ -236,30 +236,30 @@ void display() {
glBindVertexArray(sData.m_vertexArrayObject);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// glBindTexture(GL_TEXTURE_2D,sData.m_texturehandle);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
vec2 p( 0.f,0.f);//?b?0.5f * sinf(timeValue), 0.5f * cosf(timeValue) );
glUniform2fv(sData.m_positionUniform, 1, (const GLfloat *)&p);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glEnableVertexAttribArray(sData.m_positionAttribute);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glEnableVertexAttribArray(sData.m_colourAttribute);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glEnableVertexAttribArray(sData.m_textureAttribute);
@@ -267,18 +267,18 @@ void display() {
glVertexAttribPointer(sData.m_colourAttribute , 4, GL_FLOAT, GL_FALSE, sizeof(Vertex), (const GLvoid *)sizeof(vec4));
glVertexAttribPointer(sData.m_textureAttribute , 2, GL_FLOAT, GL_FALSE, sizeof(Vertex), (const GLvoid *)(sizeof(vec4)+sizeof(vec4)));
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, sData.m_indexBuffer);
//glDrawArrays(GL_TRIANGLE_FAN, 0, 4);
int indexCount = 6;
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// glDrawElements(GL_TRIANGLES, indexCount, GL_UNSIGNED_INT, 0);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// glutSwapBuffers();
}
@@ -325,8 +325,8 @@ int main(int argc, char* argv[])
printf("\n");
btgDefaultOpenGLWindow* window = new btgDefaultOpenGLWindow();
window->createWindow(btgWindowConstructionInfo(width,height));
b3gDefaultOpenGLWindow* window = new b3gDefaultOpenGLWindow();
window->createWindow(b3gWindowConstructionInfo(width,height));
window->setWindowTitle("font test");
@@ -347,7 +347,7 @@ int main(int argc, char* argv[])
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// render.InitShaders();
@@ -356,12 +356,12 @@ int main(int argc, char* argv[])
window->runMainLoop();
// window->setMouseCallback(btDefaultMouseCallback);
// window->setKeyboardCallback(btDefaultKeyboardCallback);
// window->setWheelCallback(btDefaultWheelCallback);
// window->setMouseCallback(b3DefaultMouseCallback);
// window->setKeyboardCallback(b3DefaultKeyboardCallback);
// window->setWheelCallback(b3DefaultWheelCallback);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
int done;
@@ -381,7 +381,7 @@ int main(int argc, char* argv[])
stash = sth_create(fontTextureWidth,fontTextureHeight,renderCallbacks);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
if (!stash)
{
@@ -404,9 +404,9 @@ int main(int argc, char* argv[])
#ifdef LOAD_FONT_FROM_FILE
unsigned char* data;
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
btAssert(fp);
b3Assert(fp);
if (fp)
{
fseek(fp, 0, SEEK_END);
@@ -415,7 +415,7 @@ int main(int argc, char* argv[])
data = (unsigned char*)malloc(datasize);
if (data == NULL)
{
btAssert(0);
b3Assert(0);
return -1;
}
else
@@ -425,18 +425,18 @@ int main(int argc, char* argv[])
}
if (!(droidRegular = sth_add_font_from_memory(stash, data)))
{
btAssert(0);
b3Assert(0);
return -1;
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// Load the remaining truetype fonts directly.
sprintf(fullFontFileName,"%s%s",fontPath,"DroidSerif-Italic.ttf");
if (!(droidItalic = sth_add_font(stash,fullFontFileName)))
{
btAssert(0);
b3Assert(0);
return -1;
}
sprintf(fullFontFileName,"%s%s",fontPath,"DroidSerif-Bold.ttf");
@@ -444,20 +444,20 @@ int main(int argc, char* argv[])
if (!(droidBold = sth_add_font(stash,fullFontFileName)))
{
btAssert(0);
b3Assert(0);
return -1;
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
sprintf(fullFontFileName,"%s%s",fontPath,"DroidSansJapanese.ttf");
if (!(droidJapanese = sth_add_font(stash,fullFontFileName)))
{
btAssert(0);
b3Assert(0);
return -1;
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
#else//LOAD_FONT_FROM_FILE
char* data2 = OpenSansData;
unsigned char* data = (unsigned char*) data2;
@@ -471,19 +471,19 @@ int main(int argc, char* argv[])
while (!window->requestedExit())
{
CProfileManager::Reset();
b3ProfileManager::Reset();
GLint err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
// glClearColor(0.5f,0.5f,0.5f,1.f);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
window->startRendering();
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glClearColor(1,1,1,1);//.4, .4, 0.4, 1.0);
glClear(GL_COLOR_BUFFER_BIT|GL_DEPTH_BUFFER_BIT);
@@ -491,40 +491,40 @@ int main(int argc, char* argv[])
//display();
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
if (1)
{
BT_PROFILE("font stash rendering");
B3_PROFILE("font stash rendering");
// Update and render
glEnable(GL_BLEND);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glBlendFunc(GL_SRC_ALPHA, GL_ONE_MINUS_SRC_ALPHA);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glDisable(GL_DEPTH_TEST);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
//glColor4ub(255,0,0,255);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glEnable(GL_BLEND);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
glBlendFunc(GL_SRC_ALPHA,GL_ONE_MINUS_SRC_ALPHA);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
sx = 0; sy = height;
@@ -548,7 +548,7 @@ int main(int argc, char* argv[])
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
if (0)
for (int i=0;i<1;i++)
@@ -558,7 +558,7 @@ int main(int argc, char* argv[])
{
//need to save this file as UTF-8 without signature, codepage 650001 in Visual Studio
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
//sth_draw_text(stash, droidJapanese,16.f, dx, dy-36, (const char*) "\xE7\xA7\x81\xE3\x81\xAF\xE3\x82\xAC\xE3\x83\xA9\xE3\x82\xB9\xE3\x82\x92\xE9\xA3\x9F\xE3\x81\xB9\xE3\x82\x89\xE3\x82\x8C\xE3\x81\xBE\xE3\x81\x99\xE3\x80\x82",&dx,
// width,height);//はabcdefghijlkmnopqrstuvwxyz ABCDEFGHIJKLMNOPQRSTUVWXYZ1234567890!@#$%^&*()_-+=?/\][{}.,<>`~@#$%^", &dx);
@@ -567,7 +567,7 @@ int main(int argc, char* argv[])
dx = sx;
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
sth_flush_draw(stash);
dx=0;
sth_draw_text(stash, droidRegular,14.f, dx, dy-80, "How does this OpenGL True Type font look? ", &dx,width,height);
@@ -590,7 +590,7 @@ int main(int argc, char* argv[])
sth_flush_draw(stash);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
} else
@@ -599,12 +599,12 @@ int main(int argc, char* argv[])
dy = height;
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
sth_draw_texture(stash, droidRegular, 16.f, 0, 0,width,height, "a", &dx);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
dumpTextureToPng(fontTextureWidth, fontTextureHeight,"newPic.png");
@@ -613,31 +613,31 @@ int main(int argc, char* argv[])
once++;
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
sth_end_draw(stash);
glEnable(GL_DEPTH_TEST);
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
//glFinish();
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
window->endRendering();
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
{
BT_PROFILE("glFinish");
B3_PROFILE("glFinish");
glFinish();
}
CProfileManager::Increment_Frame_Counter();
b3ProfileManager::Increment_Frame_Counter();
static bool printStats = true;
@@ -650,7 +650,7 @@ int main(int argc, char* argv[])
if (count<0)
{
count = 100;
// CProfileManager::dumpAll();
// b3ProfileManager::dumpAll();
//printStats = false;
} else
{
@@ -659,7 +659,7 @@ int main(int argc, char* argv[])
}
err = glGetError();
btAssert(err==GL_NO_ERROR);
b3Assert(err==GL_NO_ERROR);
}

2
btgui/OpenGLTrueTypeFont/stb_image_write.h
View File

@@ -276,7 +276,7 @@ unsigned char * stbi_zlib_compress(unsigned char *data, int data_len, int *out_l
stbi__sbpush(out, 0x78); // DEFLATE 32K window
stbi__sbpush(out, 0x5e); // FLEVEL = 1
stbi__zlib_add(1,1); // BFINAL = 1
stbi__zlib_add(1,2); // BTYPE = 1 -- fixed huffman
stbi__zlib_add(1,2); // B3YPE = 1 -- fixed huffman
for (i=0; i < stbi__ZHASH; ++i)
hash_table[i] = NULL;

72
btgui/OpenGLWindow/GLInstancingRenderer.cpp
View File

@@ -30,11 +30,11 @@ subject to the following restrictions:
//#include "../../opencl/gpu_rigidbody_pipeline/btGpuNarrowphaseAndSolver.h"//for m_maxNumObjectCapacity
//#include "../../opencl/gpu_rigidbody_pipeline/b3GpuNarrowphaseAndSolver.h"//for m_maxNumObjectCapacity
static InternalDataRenderer* sData2;
struct btGraphicsInstance
struct b3GraphicsInstance
{
GLuint m_cube_vao;
GLuint m_index_vbo;
@@ -49,7 +49,7 @@ struct btGraphicsInstance
int m_vertexArrayOffset;
int m_primitiveType;
btGraphicsInstance() :m_cube_vao(-1),m_index_vbo(-1),m_numIndices(-1),m_numVertices(-1),m_numGraphicsInstances(0),m_instanceOffset(0),m_vertexArrayOffset(0),m_primitiveType(BT_GL_TRIANGLES),m_texturehandle(0)
b3GraphicsInstance() :m_cube_vao(-1),m_index_vbo(-1),m_numIndices(-1),m_numVertices(-1),m_numGraphicsInstances(0),m_instanceOffset(0),m_vertexArrayOffset(0),m_primitiveType(B3_GL_TRIANGLES),m_texturehandle(0)
{
}
@@ -124,7 +124,7 @@ struct InternalDataRenderer : public GLInstanceRendererInternalData
{
if (!m_mouseButton)
{
if (btFabs(deltax)>btFabs(deltay))
if (b3Fabs(deltax)>b3Fabs(deltay))
{
m_azi -= deltax*0.1;
@@ -137,7 +137,7 @@ struct InternalDataRenderer : public GLInstanceRendererInternalData
}
} else
{
if (btFabs(deltax)>btFabs(deltay))
if (b3Fabs(deltax)>b3Fabs(deltay))
{
b3Vector3 fwd = m_cameraTargetPosition-m_cameraPosition;
b3Vector3 side = m_cameraUp.cross(fwd);
@@ -158,7 +158,7 @@ struct InternalDataRenderer : public GLInstanceRendererInternalData
{
float xDelta = x-m_mouseXpos;
float yDelta = y-m_mouseYpos;
// if (btFabs(xDelta)>btFabs(yDelta))
// if (b3Fabs(xDelta)>b3Fabs(yDelta))
// {
m_azi += xDelta*0.1;
// } else
@@ -191,23 +191,23 @@ struct GLInstanceRendererInternalData* GLInstancingRenderer::getInternalData()
return m_data;
}
void btDefaultWheelCallback(float deltax, float deltay)
void b3DefaultWheelCallback(float deltax, float deltay)
{
if (sData2)
sData2->wheelCallback(deltax,deltay);
}
void btDefaultMouseButtonCallback(int button, int state, float x, float y)
void b3DefaultMouseButtonCallback(int button, int state, float x, float y)
{
if (sData2)
sData2->mouseButtonCallback(button, state, x, y);
}
void btDefaultMouseMoveCallback( float x, float y)
void b3DefaultMouseMoveCallback( float x, float y)
{
if (sData2)
sData2->mouseMoveCallback( x, y);
}
void btDefaultKeyboardCallback(int key, int state)
void b3DefaultKeyboardCallback(int key, int state)
{
}
@@ -535,11 +535,11 @@ void GLInstancingRenderer::writeSingleInstanceTransformToGPU(float* position, fl
glFlush();
char* orgBase = (char*)glMapBuffer( GL_ARRAY_BUFFER,GL_READ_WRITE);
//btGraphicsInstance* gfxObj = m_graphicsInstances[k];
//b3GraphicsInstance* gfxObj = m_graphicsInstances[k];
int totalNumInstances= 0;
for (int k=0;k<m_graphicsInstances.size();k++)
{
btGraphicsInstance* gfxObj = m_graphicsInstances[k];
b3GraphicsInstance* gfxObj = m_graphicsInstances[k];
totalNumInstances+=gfxObj->m_numGraphicsInstances;
}
@@ -581,7 +581,7 @@ void GLInstancingRenderer::writeTransforms()
for (int k=0;k<m_graphicsInstances.size();k++)
{
btGraphicsInstance* gfxObj = m_graphicsInstances[k];
b3GraphicsInstance* gfxObj = m_graphicsInstances[k];
totalNumInstances+=gfxObj->m_numGraphicsInstances;
}
@@ -590,7 +590,7 @@ void GLInstancingRenderer::writeTransforms()
for (int k=0;k<m_graphicsInstances.size();k++)
{
//int k=0;
btGraphicsInstance* gfxObj = m_graphicsInstances[k];
b3GraphicsInstance* gfxObj = m_graphicsInstances[k];
@@ -654,10 +654,10 @@ void GLInstancingRenderer::writeTransforms()
int GLInstancingRenderer::registerGraphicsInstance(int shapeIndex, const float* position, const float* quaternion, const float* color, const float* scaling)
{
btAssert(shapeIndex == (m_graphicsInstances.size()-1));
btAssert(m_graphicsInstances.size()<m_maxNumObjectCapacity-1);
b3Assert(shapeIndex == (m_graphicsInstances.size()-1));
b3Assert(m_graphicsInstances.size()<m_maxNumObjectCapacity-1);
btGraphicsInstance* gfxObj = m_graphicsInstances[shapeIndex];
b3GraphicsInstance* gfxObj = m_graphicsInstances[shapeIndex];
int index = gfxObj->m_numGraphicsInstances + gfxObj->m_instanceOffset;
@@ -689,12 +689,12 @@ int GLInstancingRenderer::registerGraphicsInstance(int shapeIndex, const float*
int GLInstancingRenderer::registerShape(const float* vertices, int numvertices, const int* indices, int numIndices,int primitiveType)
{
btGraphicsInstance* gfxObj = new btGraphicsInstance;
b3GraphicsInstance* gfxObj = new b3GraphicsInstance;
gfxObj->m_primitiveType = primitiveType;
if (m_graphicsInstances.size())
{
btGraphicsInstance* prevObj = m_graphicsInstances[m_graphicsInstances.size()-1];
b3GraphicsInstance* prevObj = m_graphicsInstances[m_graphicsInstances.size()-1];
gfxObj->m_instanceOffset = prevObj->m_instanceOffset + prevObj->m_numGraphicsInstances;
gfxObj->m_vertexArrayOffset = prevObj->m_vertexArrayOffset + prevObj->m_numVertices;
} else
@@ -814,7 +814,7 @@ void GLInstancingRenderer::init()
assert(err==GL_NO_ERROR);
{
BT_PROFILE("texture");
B3_PROFILE("texture");
if(m_textureenabled)
{
if(!m_textureinitialized)
@@ -918,7 +918,7 @@ void GLInstancingRenderer::init()
}
void btCreateFrustum(
void b3CreateFrustum(
float left,
float right,
float bottom,
@@ -952,7 +952,7 @@ void btCreateFrustum(
void btCreateLookAt(const b3Vector3& eye, const b3Vector3& center,const b3Vector3& up, GLfloat result[16])
void b3CreateLookAt(const b3Vector3& eye, const b3Vector3& center,const b3Vector3& up, GLfloat result[16])
{
b3Vector3 f = (center - eye).normalized();
b3Vector3 u = up.normalized();
@@ -1007,7 +1007,7 @@ void GLInstancingRenderer::updateCamera()
eyePos[m_forwardAxis] = -m_data->m_cameraDistance;
b3Vector3 forward(eyePos[0],eyePos[1],eyePos[2]);
if (forward.length2() < SIMD_EPSILON)
if (forward.length2() < B3_EPSILON)
{
forward.setValue(1.f,0.f,0.f);
}
@@ -1049,13 +1049,13 @@ void GLInstancingRenderer::updateCamera()
if (m_glutScreenWidth > m_glutScreenHeight)
{
btCreateFrustum(-aspect * m_frustumZNear, aspect * m_frustumZNear, -m_frustumZNear, m_frustumZNear, m_frustumZNear, m_frustumZFar,projectionMatrix);
b3CreateFrustum(-aspect * m_frustumZNear, aspect * m_frustumZNear, -m_frustumZNear, m_frustumZNear, m_frustumZNear, m_frustumZFar,projectionMatrix);
} else
{
btCreateFrustum(-aspect * m_frustumZNear, aspect * m_frustumZNear, -m_frustumZNear, m_frustumZNear, m_frustumZNear, m_frustumZFar,projectionMatrix);
b3CreateFrustum(-aspect * m_frustumZNear, aspect * m_frustumZNear, -m_frustumZNear, m_frustumZNear, m_frustumZNear, m_frustumZFar,projectionMatrix);
}
btCreateLookAt(m_data->m_cameraPosition,m_data->m_cameraTargetPosition,m_data->m_cameraUp,modelviewMatrix);
b3CreateLookAt(m_data->m_cameraPosition,m_data->m_cameraTargetPosition,m_data->m_cameraUp,modelviewMatrix);
}
@@ -1108,7 +1108,7 @@ void GLInstancingRenderer::getMouseDirection(float* dir, int x, int y)
float bottom = -1.f;
float nearPlane = 1.f;
float tanFov = (top-bottom)*0.5f / nearPlane;
float fov = b3Scalar(2.0) * btAtan(tanFov);
float fov = b3Scalar(2.0) * b3Atan(tanFov);
b3Vector3 rayFrom = m_data->m_cameraPosition;
b3Vector3 rayForward = (m_data->m_cameraTargetPosition-m_data->m_cameraPosition);
@@ -1156,10 +1156,10 @@ void GLInstancingRenderer::getMouseDirection(float* dir, int x, int y)
void GLInstancingRenderer::RenderScene(void)
{
BT_PROFILE("GLInstancingRenderer::RenderScene");
B3_PROFILE("GLInstancingRenderer::RenderScene");
{
BT_PROFILE("init");
B3_PROFILE("init");
init();
}
@@ -1167,7 +1167,7 @@ void GLInstancingRenderer::RenderScene(void)
assert(err==GL_NO_ERROR);
{
BT_PROFILE("updateCamera");
B3_PROFILE("updateCamera");
updateCamera();
}
err = glGetError();
@@ -1200,7 +1200,7 @@ void GLInstancingRenderer::RenderScene(void)
// glBindBuffer(GL_ARRAY_BUFFER, 0);
{
BT_PROFILE("glFlush2");
B3_PROFILE("glFlush2");
glBindBuffer(GL_ARRAY_BUFFER, m_data->m_vbo);
glFlush();
@@ -1226,7 +1226,7 @@ void GLInstancingRenderer::RenderScene(void)
for (int i=0;i<m_graphicsInstances.size();i++)
{
btGraphicsInstance* gfxObj = m_graphicsInstances[i];
b3GraphicsInstance* gfxObj = m_graphicsInstances[i];
if (gfxObj->m_numGraphicsInstances)
{
// int myOffset = gfxObj->m_instanceOffset*4*sizeof(float);
@@ -1272,7 +1272,7 @@ void GLInstancingRenderer::RenderScene(void)
{
BT_PROFILE("glFlush");
B3_PROFILE("glFlush");
glFlush();
}
@@ -1281,9 +1281,9 @@ void GLInstancingRenderer::RenderScene(void)
glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, gfxObj->m_index_vbo);
{
BT_PROFILE("glDrawElementsInstanced");
B3_PROFILE("glDrawElementsInstanced");
if (gfxObj->m_primitiveType==BT_GL_POINTS)
if (gfxObj->m_primitiveType==B3_GL_POINTS)
{
glUseProgram(instancingShaderPointSprite);
glUniformMatrix4fv(ProjectionMatrixPointSprite, 1, false, &projectionMatrix[0]);
@@ -1321,7 +1321,7 @@ void GLInstancingRenderer::RenderScene(void)
err = glGetError();
assert(err==GL_NO_ERROR);
{
BT_PROFILE("glUseProgram(0);");
B3_PROFILE("glUseProgram(0);");
glUseProgram(0);
glBindBuffer(GL_ARRAY_BUFFER,0);
glBindVertexArray(0);

16
btgui/OpenGLWindow/GLInstancingRenderer.h
View File

@@ -18,21 +18,21 @@ subject to the following restrictions:
#include "Bullet3Common/b3AlignedObjectArray.h"
void btDefaultMouseButtonCallback( int button, int state, float x, float y);
void btDefaultMouseMoveCallback( float x, float y);
void btDefaultKeyboardCallback(int key, int state);
void btDefaultWheelCallback( float deltax, float deltay);
void b3DefaultMouseButtonCallback( int button, int state, float x, float y);
void b3DefaultMouseMoveCallback( float x, float y);
void b3DefaultKeyboardCallback(int key, int state);
void b3DefaultWheelCallback( float deltax, float deltay);
enum
{
BT_GL_TRIANGLES = 1,
BT_GL_POINTS
B3_GL_TRIANGLES = 1,
B3_GL_POINTS
};
class GLInstancingRenderer
{
b3AlignedObjectArray<struct btGraphicsInstance*> m_graphicsInstances;
b3AlignedObjectArray<struct b3GraphicsInstance*> m_graphicsInstances;
int m_maxNumObjectCapacity;
int m_maxShapeCapacityInBytes;
@@ -54,7 +54,7 @@ public:
void CleanupShaders();
///vertices must be in the format x,y,z, nx,ny,nz, u,v
int registerShape(const float* vertices, int numvertices, const int* indices, int numIndices, int primitiveType=BT_GL_TRIANGLES);
int registerShape(const float* vertices, int numvertices, const int* indices, int numIndices, int primitiveType=B3_GL_TRIANGLES);
///position x,y,z, quaternion x,y,z,w, color r,g,b,a, scaling x,y,z
int registerGraphicsInstance(int shapeIndex, const float* position, const float* quaternion, const float* color, const float* scaling);

32
btgui/OpenGLWindow/MacOpenGLWindow.h
View File

@@ -1,22 +1,22 @@
#ifndef MAC_OPENGL_WINDOW_H
#define MAC_OPENGL_WINDOW_H
#include "btgWindowInterface.h"
#include "b3gWindowInterface.h"
#define btgDefaultOpenGLWindow MacOpenGLWindow
#define b3gDefaultOpenGLWindow MacOpenGLWindow
class MacOpenGLWindow : public btgWindowInterface
class MacOpenGLWindow : public b3gWindowInterface
{
struct MacOpenGLWindowInternalData* m_internalData;
float m_mouseX;
float m_mouseY;
btMouseButtonCallback m_mouseButtonCallback;
btMouseMoveCallback m_mouseMoveCallback;
btWheelCallback m_wheelCallback;
btKeyboardCallback m_keyboardCallback;
btRenderCallback m_renderCallback;
b3MouseButtonCallback m_mouseButtonCallback;
b3MouseMoveCallback m_mouseMoveCallback;
b3WheelCallback m_wheelCallback;
b3KeyboardCallback m_keyboardCallback;
b3RenderCallback m_renderCallback;
float m_retinaScaleFactor;
public:
@@ -40,28 +40,28 @@ public:
void runMainLoop();
void setMouseButtonCallback(btMouseButtonCallback mouseCallback)
void setMouseButtonCallback(b3MouseButtonCallback mouseCallback)
{
m_mouseButtonCallback = mouseCallback;
}
void setMouseMoveCallback(btMouseMoveCallback mouseCallback)
void setMouseMoveCallback(b3MouseMoveCallback mouseCallback)
{
m_mouseMoveCallback = mouseCallback;
}
void setResizeCallback(btResizeCallback resizeCallback);
void setResizeCallback(b3ResizeCallback resizeCallback);
void setKeyboardCallback( btKeyboardCallback keyboardCallback)
void setKeyboardCallback( b3KeyboardCallback keyboardCallback)
{
m_keyboardCallback = keyboardCallback;
}
btKeyboardCallback getKeyboardCallback()
b3KeyboardCallback getKeyboardCallback()
{
return m_keyboardCallback;
}
void setWheelCallback (btWheelCallback wheelCallback)
void setWheelCallback (b3WheelCallback wheelCallback)
{
m_wheelCallback = wheelCallback;
}
@@ -71,13 +71,13 @@ public:
return m_retinaScaleFactor;
}
virtual void createWindow(const btgWindowConstructionInfo& ci);
virtual void createWindow(const b3gWindowConstructionInfo& ci);
virtual float getTimeInSeconds();
virtual void setRenderCallback( btRenderCallback renderCallback);
virtual void setRenderCallback( b3RenderCallback renderCallback);
virtual void setWindowTitle(const char* title);

10
btgui/OpenGLWindow/Win32InternalWindowData.h
View File

@@ -26,11 +26,11 @@ struct InternalData2
int m_mouseXpos;
int m_mouseYpos;
btWheelCallback m_wheelCallback;
btMouseMoveCallback m_mouseMoveCallback;
btMouseButtonCallback m_mouseButtonCallback;
btResizeCallback m_resizeCallback;
btKeyboardCallback m_keyboardCallback;
b3WheelCallback m_wheelCallback;
b3MouseMoveCallback m_mouseMoveCallback;
b3MouseButtonCallback m_mouseButtonCallback;
b3ResizeCallback m_resizeCallback;
b3KeyboardCallback m_keyboardCallback;

2
btgui/OpenGLWindow/Win32OpenGLWindow.cpp
View File

@@ -77,7 +77,7 @@ void Win32OpenGLWindow::disableOpenGL()
void Win32OpenGLWindow::createWindow(const btgWindowConstructionInfo& ci)
void Win32OpenGLWindow::createWindow(const b3gWindowConstructionInfo& ci)
{
Win32Window::createWindow(ci);

4
btgui/OpenGLWindow/Win32OpenGLWindow.h
View File

@@ -21,7 +21,7 @@ subject to the following restrictions:
#include "Win32Window.h"
#define btgDefaultOpenGLWindow Win32OpenGLWindow
#define b3gDefaultOpenGLWindow Win32OpenGLWindow
class Win32OpenGLWindow : public Win32Window
{
@@ -40,7 +40,7 @@ public:
virtual ~Win32OpenGLWindow();
virtual void createWindow(const btgWindowConstructionInfo& ci);
virtual void createWindow(const b3gWindowConstructionInfo& ci);
virtual void closeWindow();

56
btgui/OpenGLWindow/Win32Window.cpp
View File

@@ -57,31 +57,31 @@ int getAsciiCodeFromVirtualKeycode(int virtualKeyCode)
int keycode = 0xffffffff;
switch (virtualKeyCode)
{
case VK_F1: {keycode = BTG_F1; break;}
case VK_F2: {keycode = BTG_F2; break;}
case VK_F3: {keycode = BTG_F3; break;}
case VK_F4: {keycode = BTG_F4; break;}
case VK_F5: {keycode = BTG_F5; break;}
case VK_F6: {keycode = BTG_F6; break;}
case VK_F7: {keycode = BTG_F7; break;}
case VK_F8: {keycode = BTG_F8; break;}
case VK_F9: {keycode = BTG_F9; break;}
case VK_F10: {keycode= BTG_F10; break;}
case VK_F1: {keycode = B3G_F1; break;}
case VK_F2: {keycode = B3G_F2; break;}
case VK_F3: {keycode = B3G_F3; break;}
case VK_F4: {keycode = B3G_F4; break;}
case VK_F5: {keycode = B3G_F5; break;}
case VK_F6: {keycode = B3G_F6; break;}
case VK_F7: {keycode = B3G_F7; break;}
case VK_F8: {keycode = B3G_F8; break;}
case VK_F9: {keycode = B3G_F9; break;}
case VK_F10: {keycode= B3G_F10; break;}
case VK_SPACE: {keycode= ' '; break;}
case VK_NEXT: {keycode= BTG_PAGE_DOWN; break;}
case VK_PRIOR: {keycode= BTG_PAGE_UP; break;}
case VK_NEXT: {keycode= B3G_PAGE_DOWN; break;}
case VK_PRIOR: {keycode= B3G_PAGE_UP; break;}
case VK_INSERT: {keycode= BTG_INSERT; break;}
case VK_DELETE: {keycode= BTG_DELETE; break;}
case VK_INSERT: {keycode= B3G_INSERT; break;}
case VK_DELETE: {keycode= B3G_DELETE; break;}
case VK_END:{keycode= BTG_END; break;}
case VK_HOME:{keycode= BTG_HOME; break;}
case VK_LEFT:{keycode= BTG_LEFT_ARROW; break;}
case VK_UP:{keycode= BTG_UP_ARROW; break;}
case VK_RIGHT:{keycode= BTG_RIGHT_ARROW; break;}
case VK_DOWN:{keycode= BTG_DOWN_ARROW; break;}
case VK_END:{keycode= B3G_END; break;}
case VK_HOME:{keycode= B3G_HOME; break;}
case VK_LEFT:{keycode= B3G_LEFT_ARROW; break;}
case VK_UP:{keycode= B3G_UP_ARROW; break;}
case VK_RIGHT:{keycode= B3G_RIGHT_ARROW; break;}
case VK_DOWN:{keycode= B3G_DOWN_ARROW; break;}
default:
{
keycode = MapVirtualKey( virtualKeyCode, MAPVK_VK_TO_CHAR ) & 0x0000FFFF;
@@ -332,7 +332,7 @@ void Win32Window::setWindowTitle(const char* titleChar)
#endif
}
void Win32Window::createWindow(const btgWindowConstructionInfo& ci)
void Win32Window::createWindow(const b3gWindowConstructionInfo& ci)
{
int oglViewportWidth = ci.m_width;
int oglViewportHeight = ci.m_height;
@@ -570,7 +570,7 @@ Win32Window::~Win32Window()
}
void Win32Window::setRenderCallback( btRenderCallback renderCallback)
void Win32Window::setRenderCallback( b3RenderCallback renderCallback)
{
}
@@ -643,35 +643,35 @@ bool Win32Window::requestedExit() const
return m_data->m_quit;
}
void Win32Window::setWheelCallback(btWheelCallback wheelCallback)
void Win32Window::setWheelCallback(b3WheelCallback wheelCallback)
{
m_data->m_wheelCallback = wheelCallback;
}
void Win32Window::setMouseMoveCallback(btMouseMoveCallback mouseCallback)
void Win32Window::setMouseMoveCallback(b3MouseMoveCallback mouseCallback)
{
m_data->m_mouseMoveCallback = mouseCallback;
}
void Win32Window::setMouseButtonCallback(btMouseButtonCallback mouseCallback)
void Win32Window::setMouseButtonCallback(b3MouseButtonCallback mouseCallback)
{
m_data->m_mouseButtonCallback = mouseCallback;
}
void Win32Window::setResizeCallback(btResizeCallback resizeCallback)
void Win32Window::setResizeCallback(b3ResizeCallback resizeCallback)
{
m_data->m_resizeCallback = resizeCallback;
if (m_data->m_resizeCallback)
(*m_data->m_resizeCallback)(m_data->m_openglViewportWidth,m_data->m_openglViewportHeight);
}
void Win32Window::setKeyboardCallback( btKeyboardCallback keyboardCallback)
void Win32Window::setKeyboardCallback( b3KeyboardCallback keyboardCallback)
{
m_data->m_keyboardCallback = keyboardCallback;
}
btKeyboardCallback Win32Window::getKeyboardCallback()
b3KeyboardCallback Win32Window::getKeyboardCallback()
{
return m_data->m_keyboardCallback;
}

20
btgui/OpenGLWindow/Win32Window.h
View File

@@ -21,9 +21,9 @@ subject to the following restrictions:
struct InternalData2;
#include "btgWindowInterface.h"
#include "b3gWindowInterface.h"
class Win32Window : public btgWindowInterface
class Win32Window : public b3gWindowInterface
{
protected:
@@ -40,7 +40,7 @@ public:
virtual ~Win32Window();
virtual void createWindow(const btgWindowConstructionInfo& ci);
virtual void createWindow(const b3gWindowConstructionInfo& ci);
virtual void switchFullScreen(bool fullscreen,int width=0,int height=0,int colorBitsPerPixel=0);
@@ -64,15 +64,15 @@ public:
virtual void getMouseCoordinates(int& x, int& y);
virtual void setMouseMoveCallback(btMouseMoveCallback mouseCallback);
virtual void setMouseButtonCallback(btMouseButtonCallback mouseCallback);
virtual void setResizeCallback(btResizeCallback resizeCallback);
virtual void setWheelCallback(btWheelCallback wheelCallback);
virtual void setKeyboardCallback( btKeyboardCallback keyboardCallback);
virtual void setMouseMoveCallback(b3MouseMoveCallback mouseCallback);
virtual void setMouseButtonCallback(b3MouseButtonCallback mouseCallback);
virtual void setResizeCallback(b3ResizeCallback resizeCallback);
virtual void setWheelCallback(b3WheelCallback wheelCallback);
virtual void setKeyboardCallback( b3KeyboardCallback keyboardCallback);
virtual btKeyboardCallback getKeyboardCallback();
virtual b3KeyboardCallback getKeyboardCallback();
virtual void setRenderCallback( btRenderCallback renderCallback);
virtual void setRenderCallback( b3RenderCallback renderCallback);
virtual void setWindowTitle(const char* title);
};

56
btgui/OpenGLWindow/X11OpenGLWindow.cpp
View File

@@ -23,11 +23,11 @@ struct InternalData2
XWindowAttributes m_gwa;
XEvent m_xev;
btWheelCallback m_wheelCallback;
btMouseMoveCallback m_mouseMoveCallback;
btMouseButtonCallback m_mouseButtonCallback;
btResizeCallback m_resizeCallback;
btKeyboardCallback m_keyboardCallback;
b3WheelCallback m_wheelCallback;
b3MouseMoveCallback m_mouseMoveCallback;
b3MouseButtonCallback m_mouseButtonCallback;
b3ResizeCallback m_resizeCallback;
b3KeyboardCallback m_keyboardCallback;
InternalData2()
:m_dpy(0),
@@ -89,7 +89,7 @@ void X11OpenGLWindow::disableOpenGL()
}
void X11OpenGLWindow::createWindow(const btgWindowConstructionInfo& ci)
void X11OpenGLWindow::createWindow(const b3gWindowConstructionInfo& ci)
{
m_data->m_dpy = XOpenDisplay(NULL);
@@ -139,23 +139,23 @@ int X11OpenGLWindow::getAsciiCodeFromVirtualKeycode(int keycode)
key = XKeycodeToKeysym( m_data->m_dpy, keycode, 0 );
switch( key )
{
case XK_Escape: return BTG_ESCAPE;
case XK_Escape: return B3G_ESCAPE;
case XK_F1: return BTG_F1;
case XK_F2: return BTG_F2;
case XK_F3: return BTG_F3;
case XK_F4: return BTG_F4;
case XK_F5: return BTG_F5;
case XK_F6: return BTG_F6;
case XK_F7: return BTG_F7;
case XK_F8: return BTG_F8;
case XK_F9: return BTG_F9;
case XK_F10: return BTG_F10;
case XK_F11: return BTG_F11;
case XK_F12: return BTG_F12;
case XK_F13: return BTG_F13;
case XK_F14: return BTG_F14;
case XK_F15: return BTG_F15;
case XK_F1: return B3G_F1;
case XK_F2: return B3G_F2;
case XK_F3: return B3G_F3;
case XK_F4: return B3G_F4;
case XK_F5: return B3G_F5;
case XK_F6: return B3G_F6;
case XK_F7: return B3G_F7;
case XK_F8: return B3G_F8;
case XK_F9: return B3G_F9;
case XK_F10: return B3G_F10;
case XK_F11: return B3G_F11;
case XK_F12: return B3G_F12;
case XK_F13: return B3G_F13;
case XK_F14: return B3G_F14;
case XK_F15: return B3G_F15;
default:
// Make uppercase
XConvertCase( key, &key_lc, &key_uc );
@@ -377,7 +377,7 @@ void X11OpenGLWindow::setRequestExit()
}
void X11OpenGLWindow::setRenderCallback( btRenderCallback renderCallback)
void X11OpenGLWindow::setRenderCallback( b3RenderCallback renderCallback)
{
}
@@ -388,27 +388,27 @@ void X11OpenGLWindow::setWindowTitle(const char* title)
}
void X11OpenGLWindow::setWheelCallback(btWheelCallback wheelCallback)
void X11OpenGLWindow::setWheelCallback(b3WheelCallback wheelCallback)
{
m_data->m_wheelCallback = wheelCallback;
}
void X11OpenGLWindow::setMouseMoveCallback(btMouseMoveCallback mouseCallback)
void X11OpenGLWindow::setMouseMoveCallback(b3MouseMoveCallback mouseCallback)
{
m_data->m_mouseMoveCallback = mouseCallback;
}
void X11OpenGLWindow::setMouseButtonCallback(btMouseButtonCallback mouseCallback)
void X11OpenGLWindow::setMouseButtonCallback(b3MouseButtonCallback mouseCallback)
{
m_data->m_mouseButtonCallback = mouseCallback;
}
void X11OpenGLWindow::setResizeCallback(btResizeCallback resizeCallback)
void X11OpenGLWindow::setResizeCallback(b3ResizeCallback resizeCallback)
{
m_data->m_resizeCallback = resizeCallback;
}
void X11OpenGLWindow::setKeyboardCallback( btKeyboardCallback keyboardCallback)
void X11OpenGLWindow::setKeyboardCallback( b3KeyboardCallback keyboardCallback)
{
m_data->m_keyboardCallback = keyboardCallback;

20
btgui/OpenGLWindow/X11OpenGLWindow.h
View File

@@ -1,11 +1,11 @@
#ifndef X11_OPENGL_WINDOW_H
#define X11_OPENGL_WINDOW_H
#define btgDefaultOpenGLWindow X11OpenGLWindow
#define b3gDefaultOpenGLWindow X11OpenGLWindow
#include "btgWindowInterface.h"
#include "b3gWindowInterface.h"
class X11OpenGLWindow : public btgWindowInterface
class X11OpenGLWindow : public b3gWindowInterface
{
struct InternalData2* m_data;
@@ -27,7 +27,7 @@ public:
virtual ~X11OpenGLWindow();
virtual void createWindow(const btgWindowConstructionInfo& ci);
virtual void createWindow(const b3gWindowConstructionInfo& ci);
virtual void closeWindow();
@@ -47,13 +47,13 @@ public:
virtual void setRequestExit() ;
virtual void setMouseMoveCallback(btMouseMoveCallback mouseCallback);
virtual void setMouseButtonCallback(btMouseButtonCallback mouseCallback);
virtual void setResizeCallback(btResizeCallback resizeCallback);
virtual void setWheelCallback(btWheelCallback wheelCallback);
virtual void setKeyboardCallback( btKeyboardCallback keyboardCallback);
virtual void setMouseMoveCallback(b3MouseMoveCallback mouseCallback);
virtual void setMouseButtonCallback(b3MouseButtonCallback mouseCallback);
virtual void setResizeCallback(b3ResizeCallback resizeCallback);
virtual void setWheelCallback(b3WheelCallback wheelCallback);
virtual void setKeyboardCallback( b3KeyboardCallback keyboardCallback);
virtual void setRenderCallback( btRenderCallback renderCallback);
virtual void setRenderCallback( b3RenderCallback renderCallback);
virtual void setWindowTitle(const char* title);

109
btgui/OpenGLWindow/b3gWindowInterface.h Normal file
View File

@@ -0,0 +1,109 @@
#ifndef B3G_WINDOW_INTERFACE_H
#define B3G_WINDOW_INTERFACE_H
typedef void (*b3WheelCallback)(float deltax, float deltay);
typedef void (*b3ResizeCallback)( float width, float height);
typedef void (*b3MouseMoveCallback)( float x, float y);
typedef void (*b3MouseButtonCallback)(int button, int state, float x, float y);
typedef void (*b3KeyboardCallback)(int keycode, int state);
typedef void (*b3RenderCallback) ();
enum {
B3G_ESCAPE = 27,
B3G_F1 = 0xff00,
B3G_F2,
B3G_F3,
B3G_F4,
B3G_F5,
B3G_F6,
B3G_F7,
B3G_F8,
B3G_F9,
B3G_F10,
B3G_F11,
B3G_F12,
B3G_F13,
B3G_F14,
B3G_F15,
B3G_LEFT_ARROW,
B3G_RIGHT_ARROW,
B3G_UP_ARROW,
B3G_DOWN_ARROW,
B3G_PAGE_UP,
B3G_PAGE_DOWN,
B3G_END,
B3G_HOME,
B3G_INSERT,
B3G_DELETE
};
struct b3gWindowConstructionInfo
{
int m_width;
int m_height;
bool m_fullscreen;
int m_colorBitsPerPixel;
void* m_windowHandle;
const char* m_title;
int m_openglVersion;
b3gWindowConstructionInfo(int width=1024, int height=768)
:m_width(width),
m_height(height),
m_fullscreen(false),
m_colorBitsPerPixel(32),
m_windowHandle(0),
m_title("title"),
m_openglVersion(3)
{
}
};
class b3gWindowInterface
{
public:
virtual ~b3gWindowInterface()
{
}
virtual void createDefaultWindow(int width, int height, const char* title)
{
b3gWindowConstructionInfo ci(width,height);
ci.m_title = title;
createWindow(ci);
}
virtual void createWindow(const b3gWindowConstructionInfo& ci)=0;
virtual void closeWindow()=0;
virtual void runMainLoop()=0;
virtual float getTimeInSeconds()=0;
virtual bool requestedExit() const = 0;
virtual void setRequestExit() = 0;
virtual void startRendering()=0;
virtual void endRendering()=0;
virtual void setMouseMoveCallback(b3MouseMoveCallback mouseCallback)=0;
virtual void setMouseButtonCallback(b3MouseButtonCallback mouseCallback)=0;
virtual void setResizeCallback(b3ResizeCallback resizeCallback)=0;
virtual void setWheelCallback(b3WheelCallback wheelCallback)=0;
virtual void setKeyboardCallback( b3KeyboardCallback keyboardCallback)=0;
virtual b3KeyboardCallback getKeyboardCallback()=0;
virtual void setRenderCallback( b3RenderCallback renderCallback) = 0;
virtual void setWindowTitle(const char* title)=0;
virtual float getRetinaScale() const =0;
};
#endif //B3G_WINDOW_INTERFACE_H

109
btgui/OpenGLWindow/btgWindowInterface.h
View File

@@ -1,109 +0,0 @@
#ifndef BTG_WINDOW_INTERFACE_H
#define BTG_WINDOW_INTERFACE_H
typedef void (*btWheelCallback)(float deltax, float deltay);
typedef void (*btResizeCallback)( float width, float height);
typedef void (*btMouseMoveCallback)( float x, float y);
typedef void (*btMouseButtonCallback)(int button, int state, float x, float y);
typedef void (*btKeyboardCallback)(int keycode, int state);
typedef void (*btRenderCallback) ();
enum {
BTG_ESCAPE = 27,
BTG_F1 = 0xff00,
BTG_F2,
BTG_F3,
BTG_F4,
BTG_F5,
BTG_F6,
BTG_F7,
BTG_F8,
BTG_F9,
BTG_F10,
BTG_F11,
BTG_F12,
BTG_F13,
BTG_F14,
BTG_F15,
BTG_LEFT_ARROW,
BTG_RIGHT_ARROW,
BTG_UP_ARROW,
BTG_DOWN_ARROW,
BTG_PAGE_UP,
BTG_PAGE_DOWN,
BTG_END,
BTG_HOME,
BTG_INSERT,
BTG_DELETE
};
struct btgWindowConstructionInfo
{
int m_width;
int m_height;
bool m_fullscreen;
int m_colorBitsPerPixel;
void* m_windowHandle;
const char* m_title;
int m_openglVersion;
btgWindowConstructionInfo(int width=1024, int height=768)
:m_width(width),
m_height(height),
m_fullscreen(false),
m_colorBitsPerPixel(32),
m_windowHandle(0),
m_title("title"),
m_openglVersion(3)
{
}
};
class btgWindowInterface
{
public:
virtual ~btgWindowInterface()
{
}
virtual void createDefaultWindow(int width, int height, const char* title)
{
btgWindowConstructionInfo ci(width,height);
ci.m_title = title;
createWindow(ci);
}
virtual void createWindow(const btgWindowConstructionInfo& ci)=0;
virtual void closeWindow()=0;
virtual void runMainLoop()=0;
virtual float getTimeInSeconds()=0;
virtual bool requestedExit() const = 0;
virtual void setRequestExit() = 0;
virtual void startRendering()=0;
virtual void endRendering()=0;
virtual void setMouseMoveCallback(btMouseMoveCallback mouseCallback)=0;
virtual void setMouseButtonCallback(btMouseButtonCallback mouseCallback)=0;
virtual void setResizeCallback(btResizeCallback resizeCallback)=0;
virtual void setWheelCallback(btWheelCallback wheelCallback)=0;
virtual void setKeyboardCallback( btKeyboardCallback keyboardCallback)=0;
virtual btKeyboardCallback getKeyboardCallback()=0;
virtual void setRenderCallback( btRenderCallback renderCallback) = 0;
virtual void setWindowTitle(const char* title)=0;
virtual float getRetinaScale() const =0;
};
#endif //BTG_WINDOW_INTERFACE_H

20
btgui/OpenGLWindow/gwenWindow.cpp
View File

@@ -152,15 +152,15 @@ public:
}
float dumpRecursive(CProfileIterator* profileIterator, Gwen::Controls::TreeNode* parentNode)
float dumpRecursive(b3ProfileIterator* profileIterator, Gwen::Controls::TreeNode* parentNode)
{
profileIterator->First();
if (profileIterator->Is_Done())
return 0.f;
float accumulated_time=0,parent_time = profileIterator->Is_Root() ? CProfileManager::Get_Time_Since_Reset() : profileIterator->Get_Current_Parent_Total_Time();
float accumulated_time=0,parent_time = profileIterator->Is_Root() ? b3ProfileManager::Get_Time_Since_Reset() : profileIterator->Get_Current_Parent_Total_Time();
int i;
int frames_since_reset = CProfileManager::Get_Frame_Count_Since_Reset();
int frames_since_reset = b3ProfileManager::Get_Frame_Count_Since_Reset();
//printf("Profiling: %s (total running time: %.3f ms) ---\n", profileIterator->Get_Current_Parent_Name(), parent_time );
float totalTime = 0.f;
@@ -175,7 +175,7 @@ public:
numChildren++;
float current_total_time = profileIterator->Get_Current_Total_Time();
accumulated_time += current_total_time;
double fraction = parent_time > SIMD_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
double fraction = parent_time > B3_EPSILON ? (current_total_time / parent_time) * 100 : 0.f;
Gwen::String name(profileIterator->Get_Current_Name());
#ifdef _WIN32
@@ -214,7 +214,7 @@ public:
}
void UpdateText(CProfileIterator* profileIterator, bool idle)
void UpdateText(b3ProfileIterator* profileIterator, bool idle)
{
static bool update=true;
@@ -232,7 +232,7 @@ public:
static double time_since_reset = 0.f;
if (!idle)
{
time_since_reset = CProfileManager::Get_Time_Since_Reset();
time_since_reset = b3ProfileManager::Get_Time_Since_Reset();
}
//Gwen::UnicodeString txt = Gwen::Utility::Format( L"FEM Settings %i fps", test );
@@ -243,7 +243,7 @@ public:
double totalTime = 0;
int frames_since_reset = CProfileManager::Get_Frame_Count_Since_Reset();
int frames_since_reset = b3ProfileManager::Get_Frame_Count_Since_Reset();
profileIterator->First();
@@ -258,10 +258,10 @@ public:
#ifdef _WIN32
Gwen::UnicodeString uname = Gwen::Utility::StringToUnicode(name);
Gwen::UnicodeString txt = Gwen::Utility::Format( L"Profiling: %s total time: %.3f ms, unaccounted %.3f %% :: %.3f ms", uname.c_str(), parent_time ,
parent_time > SIMD_EPSILON ? ((parent_time - accumulated_time) / parent_time) * 100 : 0.f, parent_time - accumulated_time);
parent_time > B3_EPSILON ? ((parent_time - accumulated_time) / parent_time) * 100 : 0.f, parent_time - accumulated_time);
#else
Gwen::UnicodeString txt = Gwen::Utility::Format( L"Profiling: %s total time: %.3f ms, unaccounted %.3f %% :: %.3f ms", name, parent_time ,
parent_time > SIMD_EPSILON ? ((parent_time - accumulated_time) / parent_time) * 100 : 0.f, parent_time - accumulated_time);
parent_time > B3_EPSILON ? ((parent_time - accumulated_time) / parent_time) * 100 : 0.f, parent_time - accumulated_time);
#endif
//sprintf(blockTime,"--- Profiling: %s (total running time: %.3f ms) ---", profileIterator->Get_Current_Parent_Name(), parent_time );
//displayProfileString(xOffset,yStart,blockTime);
@@ -364,7 +364,7 @@ void resizeGUI(int width, int height)
primRenderer->setScreenSize(width,height);
}
void processProfileData(CProfileIterator* iterator, bool idle)
void processProfileData(b3ProfileIterator* iterator, bool idle)
{
if (profWindow)
{

2
btgui/OpenGLWindow/gwenWindow.h
View File

@@ -18,7 +18,7 @@ extern Gwen::Controls::Canvas* pCanvas;
class GLPrimitiveRenderer;
void setupGUI(int width, int height, sth_stash* font, float retinaScale,GLPrimitiveRenderer* primRender);
void processProfileData(class CProfileIterator* iterator, bool idle);
void processProfileData(class b3ProfileIterator* iterator, bool idle);
void resizeGUI(int width, int height);
#endif //MY_GWEN_WINDOW_H

32
btgui/OpenGLWindow/main.cpp
View File

@@ -72,7 +72,7 @@ void Usage()
void MyMouseButtonCallback(int button, int state, float x, float y)
{
//btDefaultMouseCallback(button,state,x,y);
//b3DefaultMouseCallback(button,state,x,y);
if (pCanvas)
{
@@ -106,7 +106,7 @@ void MyResizeCallback(float width, float height)
void MyMouseMoveCallback( float x, float y)
{
//btDefaultMouseCallback(button,state,x,y);
//b3DefaultMouseCallback(button,state,x,y);
static int m_lastmousepos[2] = {0,0};
static bool isInitialized = false;
@@ -278,8 +278,8 @@ int main(int argc, char* argv[])
printf("\n");
btgDefaultOpenGLWindow* window = new btgDefaultOpenGLWindow();
btgWindowConstructionInfo wci;
b3gDefaultOpenGLWindow* window = new b3gDefaultOpenGLWindow();
b3gWindowConstructionInfo wci;
wci.m_width = g_OpenGLWidth;
wci.m_height = g_OpenGLHeight;
@@ -320,8 +320,8 @@ int main(int argc, char* argv[])
window->setMouseButtonCallback(MyMouseButtonCallback);
window->setMouseMoveCallback(MyMouseMoveCallback);
window->setResizeCallback(MyResizeCallback);
window->setKeyboardCallback(btDefaultKeyboardCallback);
window->setWheelCallback(btDefaultWheelCallback);
window->setKeyboardCallback(b3DefaultKeyboardCallback);
window->setWheelCallback(b3DefaultWheelCallback);
//GLPrimitiveRenderer* pprender = new GLPrimitiveRenderer(g_OpenGLWidth,g_OpenGLHeight);
@@ -341,22 +341,22 @@ int main(int argc, char* argv[])
pCanvas->SetSize(g_OpenGLWidth,g_OpenGLHeight);
}
setupGUI(g_OpenGLWidth,g_OpenGLHeight,stash,retinaScale,primRenderer);
class CProfileIterator* m_profileIterator;
m_profileIterator = CProfileManager::Get_Iterator();
class b3ProfileIterator* m_profileIterator;
m_profileIterator = b3ProfileManager::Get_Iterator();
glClearColor(1,1,1,1);
while (!window->requestedExit())
{
CProfileManager::Reset();
b3ProfileManager::Reset();
{
BT_PROFILE("loop");
B3_PROFILE("loop");
{
BT_PROFILE("startRendering");
B3_PROFILE("startRendering");
window->startRendering();
}
render.RenderScene();
@@ -371,7 +371,7 @@ int main(int argc, char* argv[])
float dx=0;
if (1)
{
BT_PROFILE("font sth_draw_text");
B3_PROFILE("font sth_draw_text");
glEnable(GL_BLEND);
GLint err = glGetError();
@@ -400,7 +400,7 @@ int main(int argc, char* argv[])
if (1)
{
BT_PROFILE("gwen RenderCanvas");
B3_PROFILE("gwen RenderCanvas");
if (pCanvas)
{
@@ -473,7 +473,7 @@ int main(int argc, char* argv[])
CProfileManager::Increment_Frame_Counter();
b3ProfileManager::Increment_Frame_Counter();
@@ -488,10 +488,10 @@ int main(int argc, char* argv[])
{
count = 100;
{
//BT_PROFILE("processProfileData");
//B3_PROFILE("processProfileData");
processProfileData(m_profileIterator,false);
}
//CProfileManager::dumpAll();
//b3ProfileManager::dumpAll();
//printStats = false;
} else
{

4
btgui/OpenGLWindow/renderscene.cpp
View File

@@ -46,7 +46,7 @@ struct GraphicsShape
struct InstanceGroup
{
// Bullet::btCollisionShapeData* m_shape;
// Bullet::b3CollisionShapeData* m_shape;
int m_collisionShapeIndex;
// b3AlignedObjectArray<bParse::bStructHandle*> m_rigidBodies;
@@ -114,7 +114,7 @@ void createSceneProgrammatically(GLInstancingRenderer& renderer)
int tetraShapeIndex = -1;
float position[4]={0,0,0,0};
b3Quaternion born(b3Vector3(1,0,0),SIMD_PI*0.25*0.5);
b3Quaternion born(b3Vector3(1,0,0),B3_PI*0.25*0.5);
float orn[4] = {0,0,0,1};
// float rotOrn[4] = {born.getX(),born.getY(),born.getZ(),born.getW()};//

35
build/premake4.lua
View File

@@ -91,30 +91,35 @@
if not _OPTIONS["ios"] then
include "../demo/gpudemo"
include "../btgui/MidiTest"
-- include "../demo/gpudemo"
-- include "../btgui/MidiTest"
-- include "../opencl/vector_add_simplified"
-- include "../opencl/vector_add"
include "../opencl/basic_initialize"
include "../demo/gpu_initialize"
include "../opencl/parallel_primitives/host"
include "../opencl/parallel_primitives/test"
include "../opencl/parallel_primitives/benchmark"
include "../opencl/lds_bank_conflict"
-- include "../opencl/reduce"
-- include "../opencl/gpu_broadphase/test"
-- include "../opencl/gpu_narrowphase/test"
include "../btgui/Gwen"
include "../btgui/GwenOpenGLTest"
include "../btgui/OpenGLTrueTypeFont"
-- include "../btgui/OpenGLWindow"
-- include "../demo/ObjLoader"
include "../test/OpenCL/BasicInitialize"
-- include "../test/OpenCL/BroadphaseCollision"
-- include "../test/OpenCL/NarrowphaseCollision"
-- include "../test/OpenCL/ParallelPrimitives"
include "../src/Bullet3Dynamics"
include "../src/Bullet3Common"
include "../src/Bullet3Geometry"
include "../src/Bullet3Collision"
include "../src/Bullet3Serialize/Bullet2FileLoader"
include "../test/b3DynamicBvhBroadphase"
include "../src/Bullet3OpenCL"
include "../Demos3/GpuDemos"
-- include "../demo/gpu_initialize"
-- include "../opencl/lds_bank_conflict"
-- include "../opencl/reduce"
-- include "../btgui/OpenGLTrueTypeFont"
-- include "../btgui/OpenGLWindow"
-- include "../demo/ObjLoader"
-- include "../test/b3DynamicBvhBroadphase"
end

17
demo/donttouch/btCpuDynamicsWorld.cpp
View File

@@ -1,17 +0,0 @@
#include "btCpuDynamicsWorld.h"
#include "btBulletDynamicsCommon.h"
btCpuDynamicsWorld::btCpuDynamicsWorld()
:btDiscreteDynamicsWorld(
new btCollisionDispatcher(new btDefaultCollisionConfiguration()),
new b3DynamicBvhBroadphase(),new btSequentialImpulseConstraintSolver(),
new btDefaultCollisionConfiguration()//todo: remove this!
)
{
}
btCpuDynamicsWorld::~btCpuDynamicsWorld()
{
}

24
demo/donttouch/btCpuDynamicsWorld.h
View File

@@ -1,24 +0,0 @@
#ifndef BT_CPU_DYNAMICS_WORLD_H
#define BT_CPU_DYNAMICS_WORLD_H
class btDefaultCollisionConfiguration;
class btCollisionDispatcher;
struct b3DynamicBvhBroadphase;
class btSequentialImpulseConstraintSolver;
#include "BulletDynamics/Dynamics/btDiscreteDynamicsWorld.h"
class btCpuDynamicsWorld : public btDiscreteDynamicsWorld
{
public:
btCpuDynamicsWorld();
virtual ~btCpuDynamicsWorld();
};
#endif //BT_CPU_DYNAMICS_WORLD_H

109
demo/donttouch/btGpuDynamicsWorld.h
View File

@@ -1,109 +0,0 @@
#ifndef BT_GPU_DYNAMICS_WORLD_H
#define BT_GPU_DYNAMICS_WORLD_H
class b3Vector3;
class btRigidBody;
class btCollisionObject;
struct btGpuInternalData;//use this struct to avoid 'leaking' all OpenCL headers into clients code base
class CLPhysicsDemo;
#include "Bullet3Common/b3AlignedObjectArray.h"
//#include "BulletDynamics/Dynamics/btDynamicsWorld.h"
class btGpuDynamicsWorld //: public btDynamicsWorld
{
b3AlignedObjectArray<const class btCollisionShape*> m_uniqueShapes;
b3AlignedObjectArray<int> m_uniqueShapeMapping;
CLPhysicsDemo* m_gpuPhysics;
b3Vector3 m_gravity;
bool m_once;
bool initOpenCL(int preferredDeviceIndex, int preferredPlatformIndex, bool useInterop);
void exitOpenCL();
int findOrRegisterCollisionShape(const btCollisionShape* colShape);
public:
btGpuDynamicsWorld(int preferredOpenCLPlatformIndex,int preferredOpenCLDeviceIndex);
virtual ~btGpuDynamicsWorld();
virtual int stepSimulation( b3Scalar timeStep,int maxSubSteps=1, b3Scalar fixedTimeStep=b3Scalar(1.)/b3Scalar(60.));
virtual void synchronizeMotionStates()
{
btAssert(0);
}
void debugDrawWorld() {}
void setGravity(const b3Vector3& gravity);
void addRigidBody(btRigidBody* body);
void removeCollisionObject(btCollisionObject* colObj);
b3AlignedObjectArray<class btCollisionObject*>& getCollisionObjectArray();
const b3AlignedObjectArray<class btCollisionObject*>& getCollisionObjectArray() const;
virtual void addAction(btActionInterface* action)
{
btAssert(0);
}
virtual void removeAction(btActionInterface* action)
{
btAssert(0);
}
b3Vector3 getGravity () const
{
return m_gravity;
}
virtual void addRigidBody(btRigidBody* body, short group, short mask)
{
addRigidBody(body);
}
virtual void removeRigidBody(btRigidBody* body)
{
btAssert(0);
}
virtual void setConstraintSolver(btConstraintSolver* solver)
{
btAssert(0);
}
virtual btConstraintSolver* getConstraintSolver()
{
btAssert(0);
return 0;
}
virtual btDynamicsWorldType getWorldType() const
{
return BT_GPU_PHYSICS_WORLD;
}
virtual void clearForces()
{
btAssert(0);
}
};
#endif //BT_GPU_DYNAMICS_WORLD_H

129
opencl/gpu_broadphase/test/main.cpp
View File

@@ -1,129 +0,0 @@
/*
Copyright (c) 2012 Advanced Micro Devices, 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 <stdio.h>
#include "../basic_initialize/b3OpenCLUtils.h"
#include "../host/b3GpuSapBroadphase.h"
#include "Bullet3Common/b3Vector3.h"
#include "parallel_primitives/host/btFillCL.h"
#include "parallel_primitives/host/btBoundSearchCL.h"
#include "parallel_primitives/host/btRadixSort32CL.h"
#include "parallel_primitives/host/btPrefixScanCL.h"
#include "Bullet3Common/b3CommandLineArgs.h"
#include "Bullet3Common/b3MinMax.h"
int g_nPassed = 0;
int g_nFailed = 0;
bool g_testFailed = 0;
#define TEST_INIT g_testFailed = 0;
#define TEST_ASSERT(x) if( !(x) ){g_testFailed = 1;}
#define TEST_REPORT(testName) printf("[%s] %s\n",(g_testFailed)?"X":"O", testName); if(g_testFailed) g_nFailed++; else g_nPassed++;
#define NEXTMULTIPLEOF(num, alignment) (((num)/(alignment) + (((num)%(alignment)==0)?0:1))*(alignment))
cl_context g_context=0;
cl_device_id g_device=0;
cl_command_queue g_queue =0;
const char* g_deviceName = 0;
void initCL(int preferredDeviceIndex, int preferredPlatformIndex)
{
void* glCtx=0;
void* glDC = 0;
int ciErrNum = 0;
//bound search and radix sort only work on GPU right now (assume 32 or 64 width workgroup without barriers)
cl_device_type deviceType = CL_DEVICE_TYPE_ALL;
g_context = b3OpenCLUtils::createContextFromType(deviceType, &ciErrNum, 0,0,preferredDeviceIndex, preferredPlatformIndex);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numDev = b3OpenCLUtils::getNumDevices(g_context);
if (numDev>0)
{
btOpenCLDeviceInfo info;
g_device= b3OpenCLUtils::getDevice(g_context,0);
g_queue = clCreateCommandQueue(g_context, g_device, 0, &ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
b3OpenCLUtils::printDeviceInfo(g_device);
b3OpenCLUtils::getDeviceInfo(g_device,&info);
g_deviceName = info.m_deviceName;
}
}
void exitCL()
{
clReleaseCommandQueue(g_queue);
clReleaseContext(g_context);
}
inline void broadphaseTest()
{
TEST_INIT;
b3GpuSapBroadphase* sap = new b3GpuSapBroadphase(g_context,g_device,g_queue);
int group=1;
int mask=1;
b3Vector3 aabbMin(0,0,0);
b3Vector3 aabbMax(1,1,1);
int usrPtr = 1;
sap->createProxy(aabbMin,aabbMax,usrPtr,group,mask);
aabbMin.setValue(1,1,1);
aabbMax.setValue(2,2,2);
usrPtr = 2;
sap->createProxy(aabbMin,aabbMax,usrPtr,group,mask);
sap->writeAabbsToGpu();
sap->calculateOverlappingPairs();
int numOverlap = sap->getNumOverlap();
cl_mem buf = sap->getOverlappingPairBuffer();
TEST_ASSERT(numOverlap==1);
delete sap;
TEST_REPORT( "broadphaseTest" );
}
int main(int argc, char** argv)
{
int preferredDeviceIndex = -1;
int preferredPlatformIndex = -1;
b3CommandLineArgs args(argc, argv);
args.GetCmdLineArgument("deviceId", preferredDeviceIndex);
args.GetCmdLineArgument("platformId", preferredPlatformIndex);
initCL(preferredDeviceIndex,preferredPlatformIndex);
broadphaseTest();
printf("%d tests passed\n",g_nPassed, g_nFailed);
if (g_nFailed)
{
printf("%d tests failed\n",g_nFailed);
}
printf("End, press <enter>\n");
getchar();
exitCL();
}

46
opencl/gpu_broadphase/test/premake4.lua
View File

@@ -1,46 +0,0 @@
function createProject(vendor)
hasCL = findOpenCL(vendor)
if (hasCL) then
project ("OpenCL_broadphase_test_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../../bin"
includedirs {"..","../..","../../../src"}
files {
"main.cpp",
"../../basic_initialize/b3OpenCLInclude.h",
"../../basic_initialize/b3OpenCLUtils.cpp",
"../../basic_initialize/b3OpenCLUtils.h",
"../host/b3GpuSapBroadphase.cpp",
"../host/b3GpuSapBroadphase.h",
"../../parallel_primitives/host/btFillCL.cpp",
"../../parallel_primitives/host/btFillCL.h",
"../../parallel_primitives/host/btBoundSearchCL.cpp",
"../../parallel_primitives/host/btBoundSearchCL.h",
"../../parallel_primitives/host/btPrefixScanCL.cpp",
"../../parallel_primitives/host/btPrefixScanCL.h",
"../../parallel_primitives/host/btRadixSort32CL.cpp",
"../../parallel_primitives/host/btRadixSort32CL.h",
"../../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../../src/Bullet3Common/b3AlignedAllocator.h",
"../../../src/Bullet3Common/b3AlignedObjectArray.h",
"../../../src/Bullet3Common/b3Quickprof.cpp",
"../../../src/Bullet3Common/b3Quickprof.h",
}
end
end
createProject("AMD")
createProject("Intel")
createProject("NVIDIA")
createProject("Apple")

111
opencl/gpu_narrowphase/test/main.cpp
View File

@@ -1,111 +0,0 @@
/*
Copyright (c) 2012 Advanced Micro Devices, 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 <stdio.h>
#include "../basic_initialize/b3OpenCLUtils.h"
#include "../host/b3ConvexHullContact.h"
#include "Bullet3Common/b3Vector3.h"
#include "parallel_primitives/host/btFillCL.h"
#include "parallel_primitives/host/btBoundSearchCL.h"
#include "parallel_primitives/host/btRadixSort32CL.h"
#include "parallel_primitives/host/btPrefixScanCL.h"
#include "Bullet3Common/b3CommandLineArgs.h"
#include "../host/b3ConvexHullContact.h"
#include "Bullet3Common/b3MinMax.h"
int g_nPassed = 0;
int g_nFailed = 0;
bool g_testFailed = 0;
#define TEST_INIT g_testFailed = 0;
#define TEST_ASSERT(x) if( !(x) ){g_testFailed = 1;}
#define TEST_REPORT(testName) printf("[%s] %s\n",(g_testFailed)?"X":"O", testName); if(g_testFailed) g_nFailed++; else g_nPassed++;
#define NEXTMULTIPLEOF(num, alignment) (((num)/(alignment) + (((num)%(alignment)==0)?0:1))*(alignment))
cl_context g_context=0;
cl_device_id g_device=0;
cl_command_queue g_queue =0;
const char* g_deviceName = 0;
void initCL(int preferredDeviceIndex, int preferredPlatformIndex)
{
void* glCtx=0;
void* glDC = 0;
int ciErrNum = 0;
//bound search and radix sort only work on GPU right now (assume 32 or 64 width workgroup without barriers)
cl_device_type deviceType = CL_DEVICE_TYPE_ALL;
g_context = b3OpenCLUtils::createContextFromType(deviceType, &ciErrNum, 0,0,preferredDeviceIndex, preferredPlatformIndex);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numDev = b3OpenCLUtils::getNumDevices(g_context);
if (numDev>0)
{
btOpenCLDeviceInfo info;
g_device= b3OpenCLUtils::getDevice(g_context,0);
g_queue = clCreateCommandQueue(g_context, g_device, 0, &ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
b3OpenCLUtils::printDeviceInfo(g_device);
b3OpenCLUtils::getDeviceInfo(g_device,&info);
g_deviceName = info.m_deviceName;
}
}
void exitCL()
{
clReleaseCommandQueue(g_queue);
clReleaseContext(g_context);
}
inline void gpuConvexHullContactTest()
{
TEST_INIT;
TEST_ASSERT(1);
GpuSatCollision* sat = new GpuSatCollision(g_context,g_device,g_queue);
delete sat;
TEST_REPORT( "gpuConvexHullContactTest" );
}
int main(int argc, char** argv)
{
int preferredDeviceIndex = -1; int preferredPlatformIndex = -1;
b3CommandLineArgs args(argc, argv);
args.GetCmdLineArgument("deviceId", preferredDeviceIndex);
args.GetCmdLineArgument("platformId", preferredPlatformIndex);
initCL(preferredDeviceIndex,preferredPlatformIndex);
gpuConvexHullContactTest();
printf("%d tests passed\n",g_nPassed, g_nFailed);
if (g_nFailed)
{
printf("%d tests failed\n",g_nFailed);
}
printf("End, press <enter>\n");
getchar();
exitCL();
}

49
opencl/gpu_narrowphase/test/premake4.lua
View File

@@ -1,49 +0,0 @@
function createProject(vendor)
hasCL = findOpenCL(vendor)
if (hasCL) then
project ("OpenCL_sat_test_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../../bin"
includedirs {"..","../..","../../../src"}
files {
"main.cpp",
"../../basic_initialize/b3OpenCLInclude.h",
"../../basic_initialize/b3OpenCLUtils.cpp",
"../../basic_initialize/b3OpenCLUtils.h",
"../host/**.cpp",
"../host/**.h",
"../../parallel_primitives/host/btFillCL.cpp",
"../../parallel_primitives/host/btFillCL.h",
"../../parallel_primitives/host/btBoundSearchCL.cpp",
"../../parallel_primitives/host/btBoundSearchCL.h",
"../../parallel_primitives/host/btPrefixScanCL.cpp",
"../../parallel_primitives/host/btPrefixScanCL.h",
"../../parallel_primitives/host/btRadixSort32CL.cpp",
"../../parallel_primitives/host/btRadixSort32CL.h",
"../../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../../src/Bullet3Common/b3AlignedAllocator.h",
"../../../src/Bullet3Common/b3AlignedObjectArray.h",
"../../../src/Bullet3Common/b3Quickprof.cpp",
"../../../src/Bullet3Common/b3Quickprof.h",
"../../../src/Bullet3Geometry/**.cpp",
"../../../src/Bullet3Geometry/**.h",
}
end
end
createProject("AMD")
createProject("Intel")
createProject("NVIDIA")
createProject("Apple")

171
opencl/lds_bank_conflict/lds_kernels.cl
View File

@@ -1,171 +0,0 @@
#define TILE_DIM 32
#define BLOCK_ROWS 8
/*// simple copy kernel (CUDA)
// Used as reference case representing best effective bandwidth.
__global__ void copy(float *odata, const float *idata)
{
int x = blockIdx.x * TILE_DIM + threadIdx.x;
int y = blockIdx.y * TILE_DIM + threadIdx.y;
int width = gridDim.x * TILE_DIM;
for (int j = 0; j < TILE_DIM; j+= BLOCK_ROWS)
odata[(y+j)*width + x] = idata[(y+j)*width + x];
}
*/
// simple copy kernel (OpenCL)
__kernel void copyKernel(__global float* odata, __global const float* idata)
{
int x = get_group_id(0) * get_num_groups(0) + get_local_id(0);
int y = get_group_id(1) * get_num_groups(1) + get_local_id(1);
int width = get_num_groups(0) * get_local_size(0);
for (int j = 0; j < get_num_groups(1); j+= get_local_size(1))
{
odata[(y+j)*width + x] = idata[(y+j)*width + x];
}
}
/*
// copy kernel using shared memory (CUDA)
// Also used as reference case, demonstrating effect of using shared memory.
__global__ void copySharedMem(float *odata, const float *idata)
{
__shared__ float tile[TILE_DIM * TILE_DIM];
int x = blockIdx.x * TILE_DIM + threadIdx.x;
int y = blockIdx.y * TILE_DIM + threadIdx.y;
int width = gridDim.x * TILE_DIM;
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
tile[(threadIdx.y+j)*TILE_DIM + threadIdx.x] = idata[(y+j)*width + x];
__syncthreads();
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
odata[(y+j)*width + x] = tile[(threadIdx.y+j)*TILE_DIM + threadIdx.x];
}
*/
// copy kernel using shared memory (OpenCL)
// Also used as reference case, demonstrating effect of using shared memory.
__kernel void copySharedMemKernel(__global float *odata, __global const float *idata)
{
__local float tile[TILE_DIM * TILE_DIM];
int x = get_group_id(0) * get_num_groups(0) + get_local_id(0);
int y = get_group_id(1) * get_num_groups(1) + get_local_id(1);
int width = get_num_groups(0) * get_local_size(0);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
tile[(get_local_id(1)+j)*TILE_DIM + get_local_id(0)] = idata[(y+j)*width + x];
barrier(CLK_LOCAL_MEM_FENCE);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
odata[(y+j)*width + x] = tile[(get_local_id(1)+j)*TILE_DIM + get_local_id(0)];
}
/*
// naive transpose (CUDA)
// Simplest transpose; doesn't use shared memory.
// Global memory reads are coalesced but writes are not.
__global__ void transposeNaive(float *odata, const float *idata)
{
int x = blockIdx.x * TILE_DIM + threadIdx.x;
int y = blockIdx.y * TILE_DIM + threadIdx.y;
int width = gridDim.x * TILE_DIM;
for (int j = 0; j < TILE_DIM; j+= BLOCK_ROWS)
odata[x*width + (y+j)] = idata[(y+j)*width + x];
}
*/
// naive transpose (OpenCL)
// Simplest transpose; doesn't use shared memory.
// Global memory reads are coalesced but writes are not.
__kernel void transposeNaiveKernel(__global float *odata, __global const float *idata)
{
int x = get_group_id(0) * get_num_groups(0) + get_local_id(0);
int y = get_group_id(1) * get_num_groups(1) + get_local_id(1);
int width = get_num_groups(0) * get_local_size(0);
for (int j = 0; j < TILE_DIM; j+= BLOCK_ROWS)
odata[x*width + (y+j)] = idata[(y+j)*width + x];
}
/*
// coalesced transpose (CUDA)
// Uses shared memory to achieve coalesing in both reads and writes
// Tile width == #banks causes shared memory bank conflicts.
__global__ void transposeCoalesced(float *odata, const float *idata)
{
__shared__ float tile[TILE_DIM][TILE_DIM];
int x = blockIdx.x * TILE_DIM + threadIdx.x;
int y = blockIdx.y * TILE_DIM + threadIdx.y;
int width = gridDim.x * TILE_DIM;
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
tile[threadIdx.y+j][threadIdx.x] = idata[(y+j)*width + x];
__syncthreads();
x = blockIdx.y * TILE_DIM + threadIdx.x; // transpose block offset
y = blockIdx.x * TILE_DIM + threadIdx.y;
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
odata[(y+j)*width + x] = tile[threadIdx.x][threadIdx.y + j];
}
*/
// coalesced transpose (OpenCL)
// Uses shared memory to achieve coalesing in both reads and writes
// Tile width == #banks causes shared memory bank conflicts.
__kernel void transposeCoalescedKernel(__global float *odata, __global const float *idata)
{
__local float tile[TILE_DIM][TILE_DIM];
int x = get_group_id(0) * get_num_groups(0) + get_local_id(0);
int y = get_group_id(1) * get_num_groups(1) + get_local_id(1);
int width = get_num_groups(0) * get_local_size(0);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
tile[get_local_id(1)+j][get_local_id(0)] = idata[(y+j)*width + x];
barrier(CLK_LOCAL_MEM_FENCE);
x = get_group_id(1) * TILE_DIM + get_local_id(0);
y = get_group_id(0) * TILE_DIM + get_local_id(1);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
odata[(y+j)*width + x] = tile[get_local_id(0)][get_local_id(1) + j];
}
// No bank-conflict transpose (OpenCL)
// Same as transposeCoalesced except the first tile dimension is padded
// to avoid shared memory bank conflicts.
__kernel void transposeNoBankConflictsKernel(__global float *odata, __global const float *idata)
{
__local float tile[TILE_DIM][TILE_DIM+1];
int x = get_group_id(0) * get_num_groups(0) + get_local_id(0);
int y = get_group_id(1) * get_num_groups(1) + get_local_id(1);
int width = get_num_groups(0) * get_local_size(0);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
tile[get_local_id(1)+j][get_local_id(0)] = idata[(y+j)*width + x];
barrier(CLK_LOCAL_MEM_FENCE);
x = get_group_id(1) * TILE_DIM + get_local_id(0);
y = get_group_id(0) * TILE_DIM + get_local_id(1);
for (int j = 0; j < TILE_DIM; j += BLOCK_ROWS)
odata[(y+j)*width + x] = tile[get_local_id(0)][get_local_id(1) + j];
}

361
opencl/lds_bank_conflict/main.cpp
View File

@@ -1,361 +0,0 @@
//Adapted from CUDA to OpenCL by Erwin Coumans
//See http://bitbucket.org/erwincoumans/opencl_course
// Copyright 2012 NVIDIA Corporation
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
#include "b3OpenCLUtils.h"
#include "../parallel_primitives/host/btOpenCLArray.h"
#include "../parallel_primitives/host/btLauncherCL.h"
#include "Bullet3Common/b3Quickprof.h"
#include "../parallel_primitives/host/btFillCL.h"
#include "Bullet3Common/b3CommandLineArgs.h"
#include <string.h>
#include <stdio.h>
#include <assert.h>
//make sure to update the same #define in the opencl/lds_bank_conflict/lds_kernels.cl
const int TILE_DIM = 32;
const int BLOCK_ROWS = 8;
const int NUM_REPS = 100;
// Check errors and print GB/s
void postprocess(const float *ref, const float *res, int n, float ms)
{
bool passed = true;
for (int i = 0; i < n; i++)
if (res[i] != ref[i]) {
printf("\nError: at res[%d] got %f but expected %f\n", i, res[i], ref[i]);
printf("%25s\n", "*** FAILED ***");
passed = false;
break;
}
if (passed)
printf("%20.2f\n", 2 * n * sizeof(float) * 1e-6 * NUM_REPS / ms );
}
char* loadProgSource(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;
}
int main(int argc, char **argv)
{
printf("Use --deviceId=<id> or --platformId=<id> to override OpenCL device\n");
b3CommandLineArgs args(argc,argv);
const int nx = 1024;
const int ny = 1024;
const int mem_size = nx*ny*sizeof(float);
const int num_elements = nx*ny;
btClock clock;
double startEvent=0.f;
double stopEvent=0.f;
int localSizeX = TILE_DIM;
int localSizeY = BLOCK_ROWS;
int numThreadsX = (nx/TILE_DIM)*TILE_DIM;
int numThreadsY = (ny/TILE_DIM)*BLOCK_ROWS;
int gridX = numThreadsX / localSizeX;
int gridY = numThreadsY / localSizeY;
int ciErrNum = 0;
int preferred_device = -1;
int preferred_platform = -1;
args.GetCmdLineArgument("deviceId",preferred_device);
args.GetCmdLineArgument("platformId",preferred_platform);
cl_platform_id platformId=0;
cl_context ctx=0;
cl_command_queue queue=0;
cl_device_id device=0;
cl_kernel copyKernel=0;
cl_kernel copySharedMemKernel=0;
cl_kernel transposeNaiveKernel = 0;
cl_kernel transposeCoalescedKernel = 0;
cl_kernel transposeNoBankConflictsKernel= 0;
ctx = b3OpenCLUtils::createContextFromType(CL_DEVICE_TYPE_ALL, &ciErrNum,0,0,preferred_device,preferred_platform,&platformId);
b3OpenCLUtils::printPlatformInfo(platformId);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
device = b3OpenCLUtils::getDevice(ctx,0);
b3OpenCLUtils::printDeviceInfo(device);
queue = clCreateCommandQueue(ctx, device, 0, &ciErrNum);
const char* cSourceFile = "opencl/lds_bank_conflict/lds_kernels.cl";
size_t szKernelLength;
const char* cSourceCL =0;
char relativeFileName[1024];
{
const char* prefix[]={"./","../","../../","../../../","../../../../"};
int numPrefixes = sizeof(prefix)/sizeof(char*);
for (int i=0;!cSourceCL && i<numPrefixes;i++)
{
sprintf(relativeFileName,"%s%s",prefix[i],cSourceFile);
cSourceCL = loadProgSource(relativeFileName, "", &szKernelLength);
if (cSourceCL)
{
printf("Loaded program source: %s\n", relativeFileName);
}
}
}
if (!cSourceCL)
{
printf("Couldn't find file %s, exiting\n",cSourceFile);
exit(0);
}
char flags[1024]={0};
#ifdef CL_PLATFORM_INTEL
///use this flag to allow for OpenCL kernel debugging on CPU using the Intel OpenCL run-time
//sprintf(flags,"-g -s \"%s\"","C:/develop/opencl_course/opencl/lds_bank_conflict/lds_kernels.cl");
#endif//CL_PLATFORM_INTEL
copyKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,cSourceCL,"copyKernel",&ciErrNum,0,flags);
copySharedMemKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,cSourceCL,"copySharedMemKernel",&ciErrNum,0,flags);
transposeNaiveKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,cSourceCL,"transposeNaiveKernel",&ciErrNum,0,flags);
transposeCoalescedKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,cSourceCL,"transposeCoalescedKernel",&ciErrNum,0,flags);
transposeNoBankConflictsKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,cSourceCL,"transposeNoBankConflictsKernel",&ciErrNum,0,flags);
btFillCL clMemSet(ctx,device,queue);
printf("\n============================================\n");
printf("Matrix size: %d %d, Block size: %d %d, Tile size: %d %d\n",
nx, ny, TILE_DIM, BLOCK_ROWS, TILE_DIM, TILE_DIM);
float *h_idata = (float*)malloc(mem_size);
float *h_cdata = (float*)malloc(mem_size);
float *h_tdata = (float*)malloc(mem_size);
float *gold = (float*)malloc(mem_size);
btOpenCLArray<float> d_idataCL(ctx,queue);d_idataCL.resize(num_elements);
btOpenCLArray<float> d_cdataCL(ctx,queue);d_cdataCL.resize(num_elements);
btOpenCLArray<float> d_tdataCL(ctx,queue);d_tdataCL.resize(num_elements);
// check parameters and calculate execution configuration
if (nx % TILE_DIM || ny % TILE_DIM)
{
printf("nx and ny must be a multiple of TILE_DIM\n");
goto error_exit;
}
if (TILE_DIM % BLOCK_ROWS)
{
printf("TILE_DIM must be a multiple of BLOCK_ROWS\n");
goto error_exit;
}
// host
for (int j = 0; j < ny; j++)
for (int i = 0; i < nx; i++)
h_idata[j*nx + i] = j*nx + i;
// correct result for error checking
for (int j = 0; j < ny; j++)
for (int i = 0; i < nx; i++)
{
gold[j*nx + i] = h_idata[i*nx + j];
}
d_idataCL.copyFromHostPointer(h_idata,num_elements);
// events for timing
clock.reset();
float ms;
// ------------
// time kernels
// ------------
printf("%25s%25s\n", "Routine", "Bandwidth (GB/s)");
// ----
// copy
// ----
printf("%25s", "copy");
clMemSet.execute(d_cdataCL,0.f,num_elements);
{
// warm up
btLauncherCL launcher( queue, copyKernel);
launcher.setBuffer( d_cdataCL.getBufferCL());
launcher.setBuffer( d_idataCL.getBufferCL());
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
startEvent = clock.getTimeMicroseconds()/1e3;
for (int i = 0; i < NUM_REPS; i++)
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
oclCHECKERROR(ciErrNum, CL_SUCCESS);
clFinish(queue);
stopEvent = clock.getTimeMicroseconds()/1e3;
}
ms = float(stopEvent-startEvent);
d_cdataCL.copyToHostPointer(h_cdata,num_elements,0);
postprocess(h_idata, h_cdata, nx*ny, ms);
// -------------
// copySharedMem
// -------------
printf("%25s", "shared memory copy");
clMemSet.execute(d_cdataCL,0.f,num_elements);
{
btLauncherCL launcher( queue, copySharedMemKernel);
launcher.setBuffer( d_cdataCL.getBufferCL());
launcher.setBuffer( d_idataCL.getBufferCL());
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
startEvent = clock.getTimeMicroseconds()/1e3;
for (int i = 0; i < NUM_REPS; i++)
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
oclCHECKERROR(ciErrNum, CL_SUCCESS);
clFinish(queue);
stopEvent = clock.getTimeMicroseconds()/1e3;
}
ms = float(stopEvent-startEvent);
d_cdataCL.copyToHostPointer(h_cdata,num_elements,0);
postprocess(h_idata, h_cdata, nx * ny, ms);
// --------------
// transposeNaive
// --------------
printf("%25s", "naive transpose");
clMemSet.execute(d_tdataCL,0.f,num_elements);
{
// warmup
btLauncherCL launcher( queue, transposeNaiveKernel);
launcher.setBuffer( d_tdataCL.getBufferCL());
launcher.setBuffer( d_idataCL.getBufferCL());
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
startEvent = clock.getTimeMicroseconds()/1e3;
for (int i = 0; i < NUM_REPS; i++)
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
oclCHECKERROR(ciErrNum, CL_SUCCESS);
clFinish(queue);
stopEvent = clock.getTimeMicroseconds()/1e3;
}
ms = float(stopEvent-startEvent);
d_tdataCL.copyToHostPointer(h_tdata,num_elements,0);
postprocess(gold, h_tdata, nx * ny, ms);
// ------------------
// transposeCoalesced
// ------------------
printf("%25s", "coalesced transpose");
clMemSet.execute(d_tdataCL,0.f,num_elements);
{
btLauncherCL launcher( queue, transposeCoalescedKernel);
launcher.setBuffer( d_tdataCL.getBufferCL());
launcher.setBuffer( d_idataCL.getBufferCL());
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
startEvent = clock.getTimeMicroseconds()/1e3;
for (int i = 0; i < NUM_REPS; i++)
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
oclCHECKERROR(ciErrNum, CL_SUCCESS);
clFinish(queue);
stopEvent = clock.getTimeMicroseconds()/1e3;
}
ms = float(stopEvent-startEvent);
d_tdataCL.copyToHostPointer(h_tdata,num_elements,0);
postprocess(gold, h_tdata, nx * ny, ms);
// ------------------------
// transposeNoBankConflicts
// ------------------------
printf("%25s", "conflict-free transpose");
clMemSet.execute(d_tdataCL,0.f,num_elements);
{
btLauncherCL launcher( queue, transposeNoBankConflictsKernel);
launcher.setBuffer( d_tdataCL.getBufferCL());
launcher.setBuffer( d_idataCL.getBufferCL());
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
startEvent = clock.getTimeMicroseconds()/1e3;
for (int i = 0; i < NUM_REPS; i++)
launcher.launch2D(numThreadsX,numThreadsY,localSizeX,localSizeY );
oclCHECKERROR(ciErrNum, CL_SUCCESS);
clFinish(queue);
stopEvent = clock.getTimeMicroseconds()/1e3;
}
ms = float(stopEvent-startEvent);
d_tdataCL.copyToHostPointer(h_tdata,num_elements,0);
postprocess(gold, h_tdata, nx * ny, ms);
error_exit:
// cleanup
clReleaseKernel(copyKernel);
clReleaseCommandQueue(queue);
clReleaseContext(ctx);
free(h_idata);
free(h_tdata);
free(h_cdata);
free(gold);
printf("Press <enter>\n");
getchar();
}

44
opencl/lds_bank_conflict/premake4.lua
View File

@@ -1,44 +0,0 @@
function createProject (vendor)
local hasCL = findOpenCL(vendor)
if (hasCL) then
project ( "OpenCL_lds_bank_conflict_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../bin"
links {
"OpenCL_lib_parallel_primitives_host_" .. vendor
}
includedirs {
"../basic_initialize",
"../../src"
}
files {
"main.cpp",
"../basic_initialize/b3OpenCLUtils.cpp",
"../basic_initialize/b3OpenCLUtils.h",
"../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../src/Bullet3Common/b3AlignedAllocator.h",
"../../src/Bullet3Common/b3AlignedObjectArray.h",
"../../src/Bullet3Common/b3Quickprof.cpp",
"../../src/Bullet3Common/b3Quickprof.h",
}
end
end
createProject("AMD")
createProject("NVIDIA")
createProject("Intel")
createProject("Apple")

40
opencl/parallel_primitives/benchmark/premake4.lua
View File

@@ -1,40 +0,0 @@
function createProject(vendor)
hasCL = findOpenCL(vendor)
if (hasCL) then
project ("OpenCL_radixsort_benchmark_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../../bin"
includedirs {"..","../../../src"}
links {
("OpenCL_lib_parallel_primitives_host_" .. vendor)
}
files {
"test_large_problem_sorting.cpp",
"../../basic_initialize/b3OpenCLUtils.cpp",
"../../basic_initialize/b3OpenCLUtils.h",
"../host/btFillCL.cpp",
"../host/btPrefixScanCL.cpp",
"../host/btRadixSort32CL.cpp",
"../../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../../src/Bullet3Common/b3AlignedAllocator.h",
"../../../src/Bullet3Common/b3AlignedObjectArray.h",
"../../../src/Bullet3Common/b3Quickprof.cpp",
"../../../src/Bullet3Common/b3Quickprof.h",
}
end
end
createProject("AMD")
createProject("Intel")
createProject("NVIDIA")
createProject("Apple")

711
opencl/parallel_primitives/benchmark/test_large_problem_sorting.cpp
View File

@@ -1,711 +0,0 @@
/******************************************************************************
* Copyright 2010 Duane Merrill
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*
*
*
*
* AUTHORS' REQUEST:
*
* If you use|reference|benchmark this code, please cite our Technical
* Report (http://www.cs.virginia.edu/~dgm4d/papers/RadixSortTR.pdf):
*
* @TechReport{ Merrill:Sorting:2010,
* author = "Duane Merrill and Andrew Grimshaw",
* title = "Revisiting Sorting for GPGPU Stream Architectures",
* year = "2010",
* institution = "University of Virginia, Department of Computer Science",
* address = "Charlottesville, VA, USA",
* number = "CS2010-03"
* }
*
* For more information, see our Google Code project site:
* http://code.google.com/p/back40computing/
*
* Thanks!
******************************************************************************/
/******************************************************************************
* Simple test driver program for *large-problem* radix sorting.
*
* Useful for demonstrating how to integrate radix sorting into
* your application
******************************************************************************/
/******************************************************************************
* Converted from CUDA to OpenCL/DirectCompute by Erwin Coumans
******************************************************************************/
#ifdef _WIN32
#pragma warning (disable:4996)
#endif
#include <stdlib.h>
#include <stdio.h>
#include <string.h>
#include <math.h>
#include <float.h>
#include <algorithm>
#include <string>
//#include <iostream>
#include <sstream>
/**********************
*
*/
#include "../host/btRadixSort32CL.h"
#include "../../basic_initialize/b3OpenCLUtils.h"
#include "Bullet3Common/b3Quickprof.h"
cl_context g_cxMainContext;
cl_device_id g_device;
cl_command_queue g_cqCommandQueue;
/***********************
*
*/
bool g_verbose;
///Preferred OpenCL device/platform. When < 0 then no preference is used.
///Note that b3OpenCLUtils might still use the preference of using a platform vendor that matches the SDK vendor used to build the application.
///Preferred device/platform take priority over this platform-vendor match
int gPreferredDeviceId = -1;
int gPreferredPlatformId = -1;
/******************************************************************************
* Routines
******************************************************************************/
/**
* Keys-only sorting. Uses the GPU to sort the specified vector of elements for the given
* number of iterations, displaying runtime information.
*
* @param[in] num_elements
* Size in elements of the vector to sort
* @param[in] h_keys
* Vector of keys to sort
* @param[in] iterations
* Number of times to invoke the GPU sorting primitive
* @param[in] cfg
* Config
*/
template <typename K>
void TimedSort(
unsigned int num_elements,
K *h_keys,
unsigned int iterations)
{
printf("Keys only, %d iterations, %d elements\n", iterations, num_elements);
int max_elements = num_elements;
b3AlignedObjectArray<unsigned int> hostData;
hostData.resize(num_elements);
for (int i=0;i<num_elements;i++)
{
hostData[i] = h_keys[i];
}
btRadixSort32CL sorter(g_cxMainContext,g_device,g_cqCommandQueue);
btOpenCLArray<unsigned int> gpuData(g_cxMainContext,g_cqCommandQueue);
gpuData.copyFromHost(hostData);
//sorter.executeHost(gpuData);
sorter.execute(gpuData);
b3AlignedObjectArray<unsigned int> hostDataSorted;
gpuData.copyToHost(hostDataSorted);
clFinish(g_cqCommandQueue);
{
//printf("Key-values, %d iterations, %d elements", iterations, num_elements);
// Create sorting enactor
// Perform the timed number of sorting iterations
double elapsed = 0;
float duration = 0;
btClock watch;
//warm-start
gpuData.copyFromHost(hostData);
clFinish(g_cqCommandQueue);
sorter.execute(gpuData);
watch.reset();
for (int i = 0; i < iterations; i++)
{
// Move a fresh copy of the problem into device storage
gpuData.copyFromHost(hostData);
clFinish(g_cqCommandQueue);
// Start GPU timing record
double startMs = watch.getTimeMicroseconds()/1e3;
// Call the sorting API routine
sorter.execute(gpuData);
clFinish(g_cqCommandQueue);
double stopMs = watch.getTimeMicroseconds()/1e3;
duration = stopMs - startMs;
// End GPU timing record
elapsed += (double) duration;
printf("duration = %f\n", duration);
}
// Display timing information
double avg_runtime = elapsed / iterations;
// double throughput = ((double) num_elements) / avg_runtime / 1000.0 / 1000.0;
// printf(", %f GPU ms, %f x10^9 elts/sec\n", avg_runtime, throughput);
double throughput = ((double) num_elements) / avg_runtime / 1000.0 ;
printf(", %f GPU ms, %f x10^6 elts/sec\n", avg_runtime, throughput);
gpuData.copyToHost(hostData);
for (int i=0;i<num_elements;i++)
{
h_keys[i] = hostData[i];
}
}
}
/**
* Key-value sorting. Uses the GPU to sort the specified vector of elements for the given
* number of iterations, displaying runtime information.
*
* @param[in] num_elements
* Size in elements of the vector to sort
* @param[in] h_keys
* Vector of keys to sort
* @param[in,out] h_values
* Vector of values to sort
* @param[in] iterations
* Number of times to invoke the GPU sorting primitive
* @param[in] cfg
* Config
*/
template <typename K, typename V>
void TimedSort(
unsigned int num_elements,
K *h_keys,
V *h_values,
unsigned int iterations)
{
printf("Key-values, %d iterations, %d elements\n", iterations, num_elements);
int max_elements = num_elements;
b3AlignedObjectArray<btSortData> hostData;
hostData.resize(num_elements);
for (int i=0;i<num_elements;i++)
{
hostData[i].m_key = h_keys[i];
hostData[i].m_value = h_values[i];
}
btRadixSort32CL sorter(g_cxMainContext,g_device,g_cqCommandQueue);
btOpenCLArray<btSortData> gpuData(g_cxMainContext,g_cqCommandQueue);
gpuData.copyFromHost(hostData);
//sorter.executeHost(gpuData);
sorter.execute(gpuData);
b3AlignedObjectArray<btSortData> hostDataSorted;
gpuData.copyToHost(hostDataSorted);
#if 0
for (int i=0;i<num_elements;i++)
{
printf("hostData[%d].m_key = %d\n",i, hostDataSorted[i].m_key);
printf("hostData[%d].m_value = %d\n",i,hostDataSorted[i].m_value);
}
#endif
clFinish(g_cqCommandQueue);
{
//printf("Key-values, %d iterations, %d elements", iterations, num_elements);
// Create sorting enactor
// Perform the timed number of sorting iterations
double elapsed = 0;
float duration = 0;
btClock watch;
//warm-start
gpuData.copyFromHost(hostData);
sorter.execute(gpuData);
clFinish(g_cqCommandQueue);
watch.reset();
for (int i = 0; i < iterations; i++)
{
// Move a fresh copy of the problem into device storage
gpuData.copyFromHost(hostData);
clFinish(g_cqCommandQueue);
// Start GPU timing record
double startMs = watch.getTimeMicroseconds()/1e3;
// Call the sorting API routine
sorter.execute(gpuData);
clFinish(g_cqCommandQueue);
double stopMs = watch.getTimeMicroseconds()/1e3;
duration = stopMs - startMs;
// End GPU timing record
elapsed += (double) duration;
printf("duration = %f\n", duration);
}
// Display timing information
double avg_runtime = elapsed / iterations;
// double throughput = ((double) num_elements) / avg_runtime / 1000.0 / 1000.0;
// printf(", %f GPU ms, %f x10^9 elts/sec\n", avg_runtime, throughput);
double throughput = ((double) num_elements) / avg_runtime / 1000.0 ;
printf(", %f GPU ms, %f x10^6 elts/sec\n", avg_runtime, throughput);
gpuData.copyToHost(hostData);
for (int i=0;i<num_elements;i++)
{
h_keys[i] = hostData[i].m_key;
h_values[i] = hostData[i].m_value;
}
}
}
/**
* Generates random 32-bit keys.
*
* We always take the second-order byte from rand() because the higher-order
* bits returned by rand() are commonly considered more uniformly distributed
* than the lower-order bits.
*
* We can decrease the entropy level of keys by adopting the technique
* of Thearling and Smith in which keys are computed from the bitwise AND of
* multiple random samples:
*
* entropy_reduction | Effectively-unique bits per key
* -----------------------------------------------------
* -1 | 0
* 0 | 32
* 1 | 25.95
* 2 | 17.41
* 3 | 10.78
* 4 | 6.42
* ... | ...
*
*/
template <typename K>
void RandomBits(K &key, int entropy_reduction = 0, int lower_key_bits = sizeof(K) * 8)
{
const unsigned int NUM_UCHARS = (sizeof(K) + sizeof(unsigned char) - 1) / sizeof(unsigned char);
unsigned char key_bits[NUM_UCHARS];
do {
for (int j = 0; j < NUM_UCHARS; j++) {
unsigned char quarterword = 0xff;
for (int i = 0; i <= entropy_reduction; i++) {
quarterword &= (rand() >> 7);
}
key_bits[j] = quarterword;
}
if (lower_key_bits < sizeof(K) * 8) {
unsigned long long base = 0;
memcpy(&base, key_bits, sizeof(K));
base &= (1 << lower_key_bits) - 1;
memcpy(key_bits, &base, sizeof(K));
}
memcpy(&key, key_bits, sizeof(K));
} while (key != key); // avoids NaNs when generating random floating point numbers
}
/******************************************************************************
* Templated routines for printing keys/values to the console
******************************************************************************/
template<typename T>
void PrintValue(T val) {
printf("%d", val);
}
template<>
void PrintValue<float>(float val) {
printf("%f", val);
}
template<>
void PrintValue<double>(double val) {
printf("%f", val);
}
template<>
void PrintValue<unsigned char>(unsigned char val) {
printf("%u", val);
}
template<>
void PrintValue<unsigned short>(unsigned short val) {
printf("%u", val);
}
template<>
void PrintValue<unsigned int>(unsigned int val) {
printf("%u", val);
}
template<>
void PrintValue<long>(long val) {
printf("%ld", val);
}
template<>
void PrintValue<unsigned long>(unsigned long val) {
printf("%lu", val);
}
template<>
void PrintValue<long long>(long long val) {
printf("%lld", val);
}
template<>
void PrintValue<unsigned long long>(unsigned long long val) {
printf("%llu", val);
}
/**
* Compares the equivalence of two arrays
*/
template <typename T, typename SizeT>
int CompareResults(T* computed, T* reference, SizeT len, bool verbose = true)
{
printf("\n");
for (SizeT i = 0; i < len; i++) {
if (computed[i] != reference[i]) {
printf("INCORRECT: [%lu]: ", (unsigned long) i);
PrintValue<T>(computed[i]);
printf(" != ");
PrintValue<T>(reference[i]);
if (verbose) {
printf("\nresult[...");
for (size_t j = (i >= 5) ? i - 5 : 0; (j < i + 5) && (j < len); j++) {
PrintValue<T>(computed[j]);
printf(", ");
}
printf("...]");
printf("\nreference[...");
for (size_t j = (i >= 5) ? i - 5 : 0; (j < i + 5) && (j < len); j++) {
PrintValue<T>(reference[j]);
printf(", ");
}
printf("...]");
}
return 1;
}
}
printf("CORRECT\n");
return 0;
}
/**
* Creates an example sorting problem whose keys is a vector of the specified
* number of K elements, values of V elements, and then dispatches the problem
* to the GPU for the given number of iterations, displaying runtime information.
*
* @param[in] iterations
* Number of times to invoke the GPU sorting primitive
* @param[in] num_elements
* Size in elements of the vector to sort
* @param[in] cfg
* Config
*/
template<typename K, typename V>
void TestSort(
unsigned int iterations,
int num_elements,
bool keys_only)
{
// Allocate the sorting problem on the host and fill the keys with random bytes
K *h_keys = NULL;
K *h_reference_keys = NULL;
V *h_values = NULL;
h_keys = (K*) malloc(num_elements * sizeof(K));
h_reference_keys = (K*) malloc(num_elements * sizeof(K));
if (!keys_only) h_values = (V*) malloc(num_elements * sizeof(V));
// Use random bits
for (unsigned int i = 0; i < num_elements; ++i) {
RandomBits<K>(h_keys[i], 0);
//h_keys[i] = num_elements-i;
//h_keys[i] = 0xffffffffu-i;
if (!keys_only)
h_values[i] = h_keys[i];//0xffffffffu-i;
h_reference_keys[i] = h_keys[i];
}
// Run the timing test
if (keys_only) {
TimedSort<K>(num_elements, h_keys, iterations);
} else {
TimedSort<K, V>(num_elements, h_keys, h_values, iterations);
}
// cudaThreadSynchronize();
// Display sorted key data
if (g_verbose) {
printf("\n\nKeys:\n");
for (int i = 0; i < num_elements; i++) {
PrintValue<K>(h_keys[i]);
printf(", ");
}
printf("\n\n");
}
// Verify solution
std::sort(h_reference_keys, h_reference_keys + num_elements);
CompareResults<K>(h_keys, h_reference_keys, num_elements, true);
printf("\n");
fflush(stdout);
// Free our allocated host memory
if (h_keys != NULL) free(h_keys);
if (h_values != NULL) free(h_values);
}
/**
* Displays the commandline usage for this tool
*/
void Usage()
{
printf("\ntest_large_problem_sorting [--device=<device index>] [--v] [--i=<num-iterations>] [--n=<num-elements>] [--key-values] [--deviceId=<int>] [--platformId=<int>]\n");
printf("\n");
printf("\t--v\tDisplays sorted results to the console.\n");
printf("\n");
printf("\t--i\tPerforms the sorting operation <num-iterations> times\n");
printf("\t\t\ton the device. Re-copies original input each time. Default = 1\n");
printf("\n");
printf("\t--n\tThe number of elements to comprise the sample problem\n");
printf("\t\t\tDefault = 512\n");
printf("\n");
printf("\t--key-values\tSpecifies that keys are accommodated by value pairings\n");
printf("\n");
}
/******************************************************************************
* Command-line parsing
******************************************************************************/
#include <map>
#include <algorithm>
#include <string>
class b3CommandLineArgs
{
protected:
std::map<std::string, std::string> pairs;
public:
// Constructor
b3CommandLineArgs(int argc, char **argv)
{
using namespace std;
for (int i = 1; i < argc; i++)
{
string arg = argv[i];
if ((arg[0] != '-') || (arg[1] != '-')) {
continue;
}
string::size_type pos;
string key, val;
if ((pos = arg.find( '=')) == string::npos) {
key = string(arg, 2, arg.length() - 2);
val = "";
} else {
key = string(arg, 2, pos - 2);
val = string(arg, pos + 1, arg.length() - 1);
}
pairs[key] = val;
}
}
bool CheckCmdLineFlag(const char* arg_name)
{
using namespace std;
map<string, string>::iterator itr;
if ((itr = pairs.find(arg_name)) != pairs.end()) {
return true;
}
return false;
}
template <typename T>
void GetCmdLineArgument(const char *arg_name, T &val);
int ParsedArgc()
{
return pairs.size();
}
};
template <typename T>
void b3CommandLineArgs::GetCmdLineArgument(const char *arg_name, T &val)
{
using namespace std;
map<string, string>::iterator itr;
if ((itr = pairs.find(arg_name)) != pairs.end()) {
istringstream strstream(itr->second);
strstream >> val;
}
}
template <>
void b3CommandLineArgs::GetCmdLineArgument<char*>(const char* arg_name, char* &val)
{
using namespace std;
map<string, string>::iterator itr;
if ((itr = pairs.find(arg_name)) != pairs.end()) {
string s = itr->second;
val = (char*) malloc(sizeof(char) * (s.length() + 1));
strcpy(val, s.c_str());
} else {
val = NULL;
}
}
/******************************************************************************
* Main
******************************************************************************/
extern bool gDebugSkipLoadingBinary;
int main( int argc, char** argv)
{
//gDebugSkipLoadingBinary = true;
cl_int ciErrNum;
b3CommandLineArgs args(argc,argv);
args.GetCmdLineArgument("deviceId", gPreferredDeviceId);
args.GetCmdLineArgument("platformId", gPreferredPlatformId);
printf("Initialize OpenCL using btOpenCLUtils_createContextFromType\n");
cl_platform_id platformId;
g_cxMainContext = btOpenCLUtils_createContextFromType(CL_DEVICE_TYPE_ALL, &ciErrNum, 0, 0,gPreferredDeviceId,gPreferredPlatformId,&platformId);
// g_cxMainContext = btOpenCLUtils_createContextFromType(CL_DEVICE_TYPE_GPU, &ciErrNum, 0, 0,gPreferredDeviceId,gPreferredPlatformId,&platformId);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numDev = btOpenCLUtils_getNumDevices(g_cxMainContext);
if (!numDev)
{
printf("error: no OpenCL devices\n");
exit(0);
}
int result;
int devId = 0;
g_device = btOpenCLUtils_getDevice(g_cxMainContext,devId);
btOpenCLUtils_printDeviceInfo(g_device);
// create a command-queue
g_cqCommandQueue = clCreateCommandQueue(g_cxMainContext, g_device, 0, &ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
//srand(time(NULL));
srand(0); // presently deterministic
unsigned int num_elements = 8*1024*1024;//4*1024*1024;//4*1024*1024;//257;//8*524288;//2048;//512;//524288;
unsigned int iterations = 10;
bool keys_only = true;
//
// Check command line arguments
//
if (args.CheckCmdLineFlag("help"))
{
Usage();
return 0;
}
args.GetCmdLineArgument("i", iterations);
args.GetCmdLineArgument("n", num_elements);
keys_only = !args.CheckCmdLineFlag("key-values");
g_verbose = args.CheckCmdLineFlag("v");
TestSort<unsigned int, unsigned int>(
iterations,
num_elements,
keys_only);
}

19
opencl/parallel_primitives/host/btBufferInfoCL.h
View File

@@ -1,19 +0,0 @@
#ifndef BT_BUFFER_INFO_CL_H
#define BT_BUFFER_INFO_CL_H
#include "btOpenCLArray.h"
struct btBufferInfoCL
{
//btBufferInfoCL(){}
// template<typename T>
btBufferInfoCL(cl_mem buff, bool isReadOnly = false): m_clBuffer(buff), m_isReadOnly(isReadOnly){}
cl_mem m_clBuffer;
bool m_isReadOnly;
};
#endif //BT_BUFFER_INFO_CL_H

63
opencl/parallel_primitives/host/btFillCL.h
View File

@@ -1,63 +0,0 @@
#ifndef BT_FILL_CL_H
#define BT_FILL_CL_H
#include "btOpenCLArray.h"
#include "Bullet3Common/b3Scalar.h"
#include "btInt2.h"
#include "btInt4.h"
class btFillCL
{
cl_command_queue m_commandQueue;
cl_kernel m_fillKernelInt2;
cl_kernel m_fillIntKernel;
cl_kernel m_fillUnsignedIntKernel;
cl_kernel m_fillFloatKernel;
public:
struct btConstData
{
union
{
btInt4 m_data;
btUnsignedInt4 m_UnsignedData;
};
int m_offset;
int m_n;
int m_padding[2];
};
protected:
public:
btFillCL(cl_context ctx, cl_device_id device, cl_command_queue queue);
virtual ~btFillCL();
void execute(btOpenCLArray<unsigned int>& src, const unsigned int value, int n, int offset = 0);
void execute(btOpenCLArray<int>& src, const int value, int n, int offset = 0);
void execute(btOpenCLArray<float>& src, const float value, int n, int offset = 0);
void execute(btOpenCLArray<btInt2>& src, const btInt2& value, int n, int offset = 0);
void executeHost(b3AlignedObjectArray<btInt2> &src, const btInt2 &value, int n, int offset);
void executeHost(b3AlignedObjectArray<int> &src, const int value, int n, int offset);
// void execute(btOpenCLArray<btInt4>& src, const btInt4& value, int n, int offset = 0);
};
#endif //BT_FILL_CL_H

35
opencl/parallel_primitives/host/btInt2.h
View File

@@ -1,35 +0,0 @@
#ifndef BT_INT2_H
#define BT_INT2_H
struct btUnsignedInt2
{
union
{
struct
{
unsigned int x,y;
};
struct
{
unsigned int s[2];
};
};
};
struct btInt2
{
union
{
struct
{
int x,y;
};
struct
{
int s[2];
};
};
};
#endif

55
opencl/parallel_primitives/host/btInt4.h
View File

@@ -1,55 +0,0 @@
#ifndef BT_INT4_H
#define BT_INT4_H
#include "Bullet3Common/b3Scalar.h"
ATTRIBUTE_ALIGNED16(struct) btUnsignedInt4
{
BT_DECLARE_ALIGNED_ALLOCATOR();
union
{
struct
{
unsigned int x,y,z,w;
};
struct
{
unsigned int s[4];
};
};
};
ATTRIBUTE_ALIGNED16(struct) btInt4
{
BT_DECLARE_ALIGNED_ALLOCATOR();
union
{
struct
{
int x,y,z,w;
};
struct
{
int s[4];
};
};
};
SIMD_FORCE_INLINE btInt4 btMakeInt4(int x, int y, int z, int w = 0)
{
btInt4 v;
v.s[0] = x; v.s[1] = y; v.s[2] = z; v.s[3] = w;
return v;
}
SIMD_FORCE_INLINE btUnsignedInt4 btMakeUnsignedInt4(unsigned int x, unsigned int y, unsigned int z, unsigned int w = 0)
{
btUnsignedInt4 v;
v.s[0] = x; v.s[1] = y; v.s[2] = z; v.s[3] = w;
return v;
}
#endif //BT_INT4_H

85
opencl/parallel_primitives/host/btRadixSort32CL.h
View File

@@ -1,85 +0,0 @@
#ifndef BT_RADIXSORT32_H
#define BT_RADIXSORT32_H
#include "btOpenCLArray.h"
struct btSortData
{
int m_key;
int m_value;
};
#include "btBufferInfoCL.h"
class btRadixSort32CL
{
btOpenCLArray<unsigned int>* m_workBuffer1;
btOpenCLArray<unsigned int>* m_workBuffer2;
btOpenCLArray<btSortData>* m_workBuffer3;
btOpenCLArray<btSortData>* m_workBuffer4;
btOpenCLArray<unsigned int>* m_workBuffer3a;
btOpenCLArray<unsigned int>* m_workBuffer4a;
cl_command_queue m_commandQueue;
cl_kernel m_streamCountSortDataKernel;
cl_kernel m_streamCountKernel;
cl_kernel m_prefixScanKernel;
cl_kernel m_sortAndScatterSortDataKernel;
cl_kernel m_sortAndScatterKernel;
bool m_deviceCPU;
class btPrefixScanCL* m_scan;
class btFillCL* m_fill;
public:
struct btConstData
{
int m_n;
int m_nWGs;
int m_startBit;
int m_nBlocksPerWG;
};
enum
{
DATA_ALIGNMENT = 256,
WG_SIZE = 64,
BLOCK_SIZE = 256,
ELEMENTS_PER_WORK_ITEM = (BLOCK_SIZE/WG_SIZE),
BITS_PER_PASS = 4,
NUM_BUCKET=(1<<BITS_PER_PASS),
// if you change this, change nPerWI in kernel as well
NUM_WGS = 20*6, // cypress
// NUM_WGS = 24*6, // cayman
// NUM_WGS = 32*4, // nv
};
private:
public:
btRadixSort32CL(cl_context ctx, cl_device_id device, cl_command_queue queue, int initialCapacity =0);
virtual ~btRadixSort32CL();
void execute(btOpenCLArray<unsigned int>& keysIn, btOpenCLArray<unsigned int>& keysOut, btOpenCLArray<unsigned int>& valuesIn,
btOpenCLArray<unsigned int>& valuesOut, int n, int sortBits = 32);
///keys only
void execute(btOpenCLArray<unsigned int>& keysInOut, int sortBits = 32 );
void execute(btOpenCLArray<btSortData>& keyValuesInOut, int sortBits = 32 );
void executeHost(btOpenCLArray<btSortData>& keyValuesInOut, int sortBits = 32);
void executeHost(b3AlignedObjectArray<btSortData>& keyValuesInOut, int sortBits = 32);
};
#endif //BT_RADIXSORT32_H

379
opencl/parallel_primitives/test/main.cpp
View File

@@ -1,379 +0,0 @@
/*
Copyright (c) 2012 Advanced Micro Devices, 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 <stdio.h>
#include "../basic_initialize/b3OpenCLUtils.h"
#include "../host/btFillCL.h"
#include "../host/btBoundSearchCL.h"
#include "../host/btRadixSort32CL.h"
#include "../host/btPrefixScanCL.h"
#include "Bullet3Common/b3CommandLineArgs.h"
#include "Bullet3Common/b3MinMax.h"
int g_nPassed = 0;
int g_nFailed = 0;
bool g_testFailed = 0;
#define TEST_INIT g_testFailed = 0;
#define TEST_ASSERT(x) if( !(x) ){g_testFailed = 1;}
#define TEST_REPORT(testName) printf("[%s] %s\n",(g_testFailed)?"X":"O", testName); if(g_testFailed) g_nFailed++; else g_nPassed++;
#define NEXTMULTIPLEOF(num, alignment) (((num)/(alignment) + (((num)%(alignment)==0)?0:1))*(alignment))
cl_context g_context=0;
cl_device_id g_device=0;
cl_command_queue g_queue =0;
const char* g_deviceName = 0;
void initCL(int preferredDeviceIndex, int preferredPlatformIndex)
{
void* glCtx=0;
void* glDC = 0;
int ciErrNum = 0;
//bound search and radix sort only work on GPU right now (assume 32 or 64 width workgroup without barriers)
cl_device_type deviceType = CL_DEVICE_TYPE_ALL;
g_context = b3OpenCLUtils::createContextFromType(deviceType, &ciErrNum, 0,0,preferredDeviceIndex, preferredPlatformIndex);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numDev = b3OpenCLUtils::getNumDevices(g_context);
if (numDev>0)
{
btOpenCLDeviceInfo info;
g_device= b3OpenCLUtils::getDevice(g_context,0);
g_queue = clCreateCommandQueue(g_context, g_device, 0, &ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
b3OpenCLUtils::printDeviceInfo(g_device);
b3OpenCLUtils::getDeviceInfo(g_device,&info);
g_deviceName = info.m_deviceName;
}
}
void exitCL()
{
clReleaseCommandQueue(g_queue);
clReleaseContext(g_context);
}
inline void fillIntTest()
{
TEST_INIT;
btFillCL* fillCL = new btFillCL(g_context,g_device,g_queue);
int maxSize=1024*256;
btOpenCLArray<int> intBuffer(g_context,g_queue,maxSize);
intBuffer.resize(maxSize);
#define NUM_TESTS 7
int dx = maxSize/NUM_TESTS;
for (int iter=0;iter<NUM_TESTS;iter++)
{
int size = btMin( 11+dx*iter, maxSize );
int value = 2;
int offset=0;
fillCL->execute(intBuffer,value,size,offset);
b3AlignedObjectArray<int> hostBuf2;
hostBuf2.resize(size);
fillCL->executeHost(hostBuf2,value,size,offset);
b3AlignedObjectArray<int> hostBuf;
intBuffer.copyToHost(hostBuf);
for(int i=0; i<size; i++)
{
TEST_ASSERT( hostBuf[i] == hostBuf2[i] );
TEST_ASSERT( hostBuf[i] == hostBuf2[i] );
}
}
delete fillCL;
TEST_REPORT( "fillIntTest" );
}
__inline
void seedRandom(int seed)
{
srand( seed );
}
template<typename T>
__inline
T getRandom(const T& minV, const T& maxV)
{
float r = (rand()%10000)/10000.f;
T range = maxV - minV;
return (T)(minV + r*range);
}
struct btSortDataCompare
{
inline bool operator()(const btSortData& first, const btSortData& second) const
{
return (first.m_key < second.m_key) || (first.m_key==second.m_key && first.m_value < second.m_value);
}
};
void boundSearchTest( )
{
TEST_INIT;
int maxSize = 1024*256;
int bucketSize = 256;
btOpenCLArray<btSortData> srcCL(g_context,g_queue,maxSize);
btOpenCLArray<unsigned int> upperCL(g_context,g_queue,maxSize);
btOpenCLArray<unsigned int> lowerCL(g_context,g_queue,maxSize);
b3AlignedObjectArray<btSortData> srcHost;
b3AlignedObjectArray<unsigned int> upperHost;
b3AlignedObjectArray<unsigned int> lowerHost;
b3AlignedObjectArray<unsigned int> upperHostCompare;
b3AlignedObjectArray<unsigned int> lowerHostCompare;
btBoundSearchCL* search = new btBoundSearchCL(g_context,g_device,g_queue, maxSize);
int dx = maxSize/NUM_TESTS;
for(int iter=0; iter<NUM_TESTS; iter++)
{
int size = btMin( 128+dx*iter, maxSize );
upperHost.resize(bucketSize);
lowerHost.resize(bucketSize);
upperHostCompare.resize(bucketSize);
lowerHostCompare.resize(bucketSize);
srcHost.resize(size);
for(int i=0; i<size; i++)
{
btSortData v;
// v.m_key = i<2? 0 : 5;
v.m_key = getRandom(0,bucketSize);
v.m_value = i;
srcHost.at(i) = v;
}
srcHost.quickSort(btSortDataCompare());
srcCL.copyFromHost(srcHost);
{
for(int i=0; i<bucketSize; i++)
{
lowerHost[i] = -1;
lowerHostCompare[i] = -1;
upperHost[i] = -1;
upperHostCompare[i] = -1;
}
upperCL.copyFromHost(upperHost);
lowerCL.copyFromHost(lowerHost);
}
search->execute(srcCL,size,upperCL,bucketSize,btBoundSearchCL::BOUND_UPPER);
search->execute(srcCL,size,lowerCL,bucketSize,btBoundSearchCL::BOUND_LOWER);
search->executeHost(srcHost,size,upperHostCompare,bucketSize,btBoundSearchCL::BOUND_UPPER);
search->executeHost(srcHost,size,lowerHostCompare,bucketSize,btBoundSearchCL::BOUND_LOWER);
lowerCL.copyToHost(lowerHost);
upperCL.copyToHost(upperHost);
for(int i=0; i<bucketSize; i++)
{
TEST_ASSERT(upperHostCompare[i] == upperHost[i]);
TEST_ASSERT(lowerHostCompare[i] == lowerHost[i]);
}
/*
for(int i=1; i<bucketSize; i++)
{
int lhi_1 = lowerHost[i-1];
int lhi = lowerHost[i];
for(int j=lhi_1; j<lhi; j++)
//for(int j=lowerHost[i-1]; j<lowerHost[i]; j++)
{
TEST_ASSERT( srcHost[j].m_key < i );
}
}
for(int i=0; i<bucketSize; i++)
{
int jMin = (i==0)?0:upperHost[i-1];
for(int j=jMin; j<upperHost[i]; j++)
{
TEST_ASSERT( srcHost[j].m_key <= i );
}
}
*/
for(int i=0; i<bucketSize; i++)
{
int lhi = lowerHost[i];
int uhi = upperHost[i];
for(int j=lhi; j<uhi; j++)
{
if ( srcHost[j].m_key != i )
{
printf("error %d != %d\n",srcHost[j].m_key,i);
}
TEST_ASSERT( srcHost[j].m_key == i );
}
}
}
delete search;
TEST_REPORT( "boundSearchTest" );
}
void prefixScanTest()
{
TEST_INIT;
int maxSize = 1024*256;
b3AlignedObjectArray<unsigned int> buf0Host;
b3AlignedObjectArray<unsigned int> buf1Host;
btOpenCLArray<unsigned int> buf2CL(g_context,g_queue,maxSize);
btOpenCLArray<unsigned int> buf3CL(g_context,g_queue,maxSize);
btPrefixScanCL* scan = new btPrefixScanCL(g_context,g_device,g_queue,maxSize);
int dx = maxSize/NUM_TESTS;
for(int iter=0; iter<NUM_TESTS; iter++)
{
int size = btMin( 128+dx*iter, maxSize );
buf0Host.resize(size);
buf1Host.resize(size);
for(int i=0; i<size; i++)
buf0Host[i] = 1;
buf2CL.copyFromHost( buf0Host);
unsigned int sumHost, sumGPU;
scan->executeHost(buf0Host, buf1Host, size, &sumHost );
scan->execute( buf2CL, buf3CL, size, &sumGPU );
buf3CL.copyToHost(buf0Host);
TEST_ASSERT( sumHost == sumGPU );
for(int i=0; i<size; i++)
TEST_ASSERT( buf1Host[i] == buf0Host[i] );
}
delete scan;
TEST_REPORT( "scanTest" );
}
bool radixSortTest()
{
TEST_INIT;
int maxSize = 1024*256;
b3AlignedObjectArray<btSortData> buf0Host;
buf0Host.resize(maxSize);
b3AlignedObjectArray<btSortData> buf1Host;
buf1Host.resize(maxSize );
btOpenCLArray<btSortData> buf2CL(g_context,g_queue,maxSize);
btRadixSort32CL* sort = new btRadixSort32CL(g_context,g_device,g_queue,maxSize);
int dx = maxSize/NUM_TESTS;
for(int iter=0; iter<NUM_TESTS; iter++)
{
int size = btMin( 128+dx*iter, maxSize-512 );
size = NEXTMULTIPLEOF( size, 512 );//not necessary
buf0Host.resize(size);
for(int i=0; i<size; i++)
{
btSortData v;
v.m_key = getRandom(0,0xff);
v.m_value = i;
buf0Host[i] = v;
}
buf2CL.copyFromHost( buf0Host);
sort->executeHost( buf0Host);
sort->execute(buf2CL);
buf2CL.copyToHost(buf1Host);
for(int i=0; i<size; i++)
{
TEST_ASSERT( buf0Host[i].m_value == buf1Host[i].m_value && buf0Host[i].m_key == buf1Host[i].m_key );
}
}
delete sort;
TEST_REPORT( "radixSort" );
return g_testFailed;
}
int main(int argc, char** argv)
{
int preferredDeviceIndex = -1;
int preferredPlatformIndex = -1;
b3CommandLineArgs args(argc, argv);
args.GetCmdLineArgument("deviceId", preferredDeviceIndex);
args.GetCmdLineArgument("platformId", preferredPlatformIndex);
initCL(preferredDeviceIndex,preferredPlatformIndex);
fillIntTest();
boundSearchTest();
prefixScanTest();
radixSortTest();
exitCL();
printf("%d tests passed, %d tests failed\n",g_nPassed, g_nFailed);
printf("End, press <enter>\n");
getchar();
}

41
opencl/parallel_primitives/test/premake4.lua
View File

@@ -1,41 +0,0 @@
function createProject(vendor)
hasCL = findOpenCL(vendor)
if (hasCL) then
project ("OpenCL_primitives_test_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../../bin"
includedirs {".","..","../../../src"}
files {
"main.cpp",
"../../basic_initialize/b3OpenCLInclude.h",
"../../basic_initialize/b3OpenCLUtils.cpp",
"../../basic_initialize/b3OpenCLUtils.h",
"../host/btFillCL.cpp",
"../host/btFillCL.h",
"../host/btBoundSearchCL.cpp",
"../host/btBoundSearchCL.h",
"../host/btPrefixScanCL.cpp",
"../host/btPrefixScanCL.h",
"../host/btRadixSort32CL.cpp",
"../host/btRadixSort32CL.h",
"../../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../../src/Bullet3Common/b3AlignedAllocator.h",
"../../../src/Bullet3Common/b3AlignedObjectArray.h",
}
end
end
createProject("AMD")
createProject("Intel")
createProject("NVIDIA")
createProject("Apple")

116
opencl/reduce/main.cpp
View File

@@ -1,116 +0,0 @@
///original author: Erwin Coumans
#include "b3OpenCLUtils.h"
#include "../parallel_primitives/host/btOpenCLArray.h"
#include "../parallel_primitives/host/btLauncherCL.h"
#include <stdio.h>
#define MSTRINGIFY(A) #A
const char* kernelString= MSTRINGIFY(
__kernel void ReduceGlobal(__global int* d_in, __global int* d_out, int numElements)
{
int myId = get_global_id(0);
int tid = get_local_id(0);
int ls = get_local_size(0);
for (unsigned int s=ls/2;s>0;s>>=1)
{
if (myId<numElements)
{
if (tid<s)
{
d_in[myId] += d_in[myId+s];
}
}
barrier(CLK_GLOBAL_MEM_FENCE);
}
if (tid==0)
{
if (myId<numElements)
{
d_out[get_group_id(0)]=d_in[myId];
}
}
}
);
int main(int argc, char* argv[])
{
int ciErrNum = 0;
int preferred_device = -1;
int preferred_platform = -1;
cl_platform_id platformId;
cl_context ctx;
cl_command_queue queue;
cl_device_id device;
cl_kernel addKernel;
ctx = b3OpenCLUtils::createContextFromType(CL_DEVICE_TYPE_ALL, &ciErrNum,0,0,preferred_device,preferred_platform,&platformId);
b3OpenCLUtils::printPlatformInfo(platformId);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
if (!ctx) {
printf("No OpenCL capable GPU found!");
return 0;
}
device = b3OpenCLUtils::getDevice(ctx,0);
queue = clCreateCommandQueue(ctx, device, 0, &ciErrNum);
addKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,kernelString,"ReduceGlobal",&ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numElements = 1024*1024;
btOpenCLArray<int> a(ctx,queue);
btOpenCLArray<int> b(ctx,queue);
b3AlignedObjectArray<int> hostA;
b3AlignedObjectArray<int> hostB;
for (int i=0;i<numElements;i++)
{
hostA.push_back(1);
hostB.push_back(0.f);
}
a.copyFromHost(hostA);
b.copyFromHost(hostB);
int hostSum= 0;
for (int i=0;i<numElements;i++)
{
hostSum += hostA.at(i);
}
b.resize(numElements);
{
btLauncherCL launcher( queue, addKernel);
launcher.setBuffer( a.getBufferCL());
launcher.setBuffer( b.getBufferCL());
launcher.setConst( numElements );
launcher.launch1D( numElements,1024);
}
clFinish(queue);
{
btLauncherCL launcher( queue, addKernel);
launcher.setBuffer( b.getBufferCL());
launcher.setBuffer( a.getBufferCL());
launcher.setConst( 1024 );
launcher.launch1D( 1024,1024);
}
clFinish(queue);
printf("hostSum = %d\n", hostSum);
int clSum = a.at(0);
printf("clSum = %d\n", clSum );
if (hostSum != clSum)
{
printf("Incorrect result\n");
} else
{
printf("Correct result\n");
}
clReleaseCommandQueue(queue);
clReleaseContext(ctx);
printf("press key\n");
getchar();
return 0;
}

41
opencl/reduce/premake4.lua
View File

@@ -1,41 +0,0 @@
function createProject (vendor)
local hasCL = findOpenCL(vendor)
if (hasCL) then
project ( "OpenCL_reduce_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../bin"
links {
"OpenCL_lib_parallel_primitives_host_" .. vendor
}
includedirs {
"../basic_initialize",
"../../src"
}
files {
"main.cpp",
"../basic_initialize/b3OpenCLUtils.cpp",
"../basic_initialize/b3OpenCLUtils.h",
"../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../src/Bullet3Common/b3AlignedAllocator.h",
"../../src/Bullet3Common/b3AlignedObjectArray.h",
}
end
end
createProject("AMD")
createProject("NVIDIA")
createProject("Intel")
createProject("Apple")

16
opencl/vector_add/VectorAddKernels.cl
View File

@@ -1,16 +0,0 @@
__kernel void VectorAdd(__global const float8* a, __global const float8* b, __global float8* c, int numElements)
{
// get oct-float index into global data array
int iGID = get_global_id(0);
if (iGID>=numElements)
return;
float8 aGID = a[iGID];
float8 bGID = b[iGID];
float8 result = aGID + bGID;
// write back out to GMEM
c[iGID] = result;
}

20
opencl/vector_add/VectorAddKernels.h
View File

@@ -1,20 +0,0 @@
//this file is autogenerated using stringify.bat (premake --stringify) in the build folder of this project
static const char* vectorAddCL= \
"\n"
"\n"
"__kernel void VectorAdd(__global const float8* a, __global const float8* b, __global float8* c, int numElements)\n"
"{\n"
" // get oct-float index into global data array\n"
" int iGID = get_global_id(0);\n"
" if (iGID>=numElements)\n"
" return;\n"
"\n"
" float8 aGID = a[iGID];\n"
" float8 bGID = b[iGID];\n"
"\n"
" float8 result = aGID + bGID;\n"
" // write back out to GMEM\n"
" c[iGID] = result;\n"
"}\n"
"\n"
;

408
opencl/vector_add/main.cpp
View File

@@ -1,408 +0,0 @@
///VectorAdd sample, from the NVidia JumpStart Guide
///http://developer.download.nvidia.com/OpenCL/NVIDIA_OpenCL_JumpStart_Guide.pdf
///Instead of #include <CL/cl.h> we include <MiniCL/cl.h>
///Apart from this include file, all other code should compile and work on OpenCL compliant implementation
#define LOAD_FROM_FILE
#ifdef __APPLE__
#include <OpenCL/OpenCL.h>
#else
#include <CL/cl.h>
#endif //__APPLE__
#ifdef _WIN32
#pragma warning (disable:4996)
#endif
#include <stdio.h>
#include <math.h>
#include <stdlib.h>
#include <string.h>
#define GRID3DOCL_CHECKERROR(a, b) if((a)!=(b)) { printf("3D GRID OCL Error : %d\n", (a)); btAssert((a) == (b)); }
size_t wgSize;
#include "VectorAddKernels.h"
#ifdef CL_PLATFORM_INTEL
const char* preferredPlatform = "Intel(R) Corporation";
#elif defined CL_PLATFORM_AMD
const char* preferredPlatform = "Advanced Micro Devices, Inc.";
#elif defined CL_PLATFORM_NVIDIA
const char* preferredPlatform = "NVIDIA Corporation";
#else
const char* preferredPlatform = "Unknown";
#endif
char* loadProgSource(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;
}
size_t workitem_size[3];
void printDevInfo(cl_device_id device)
{
char device_string[1024];
clGetDeviceInfo(device, CL_DEVICE_NAME, sizeof(device_string), &device_string, NULL);
printf( " Device %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 )
printf(" CL_DEVICE_TYPE:\t\t%s\n", "CL_DEVICE_TYPE_CPU");
if( type & CL_DEVICE_TYPE_GPU )
printf( " CL_DEVICE_TYPE:\t\t%s\n", "CL_DEVICE_TYPE_GPU");
if( type & CL_DEVICE_TYPE_ACCELERATOR )
printf( " CL_DEVICE_TYPE:\t\t%s\n", "CL_DEVICE_TYPE_ACCELERATOR");
if( type & CL_DEVICE_TYPE_DEFAULT )
printf( " CL_DEVICE_TYPE:\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);
printf( " CL_DEVICE_MAX_COMPUTE_UNITS:\t%d\n", compute_units);
// CL_DEVICE_MAX_WORK_GROUP_SIZE
clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(workitem_size), &workitem_size, NULL);
printf( " CL_DEVICE_MAX_WORK_ITEM_SIZES:\t%u / %u / %u \n", workitem_size[0], workitem_size[1], workitem_size[2]);
}
// Main function
// *********************************************************************
int main(int argc, char **argv)
{
void *srcA, *srcB, *dst; // Host buffers for OpenCL test
cl_context cxGPUContext; // OpenCL context
cl_command_queue cqCommandQue; // OpenCL command que
cl_device_id* cdDevices; // OpenCL device list
cl_program cpProgram; // OpenCL program
cl_kernel ckKernel; // OpenCL kernel
cl_mem cmMemObjs[3]; // OpenCL memory buffer objects: 3 for device
size_t szGlobalWorkSize[1]; // 1D var for Total # of work items
size_t szLocalWorkSize[1]; // 1D var for # of work items in the work group
size_t szParmDataBytes; // Byte size of context information
cl_int ciErr1, ciErr2; // Error code var
int iTestN = 100000 * 8; // Size of Vectors to process
int actualGlobalSize = iTestN / 8;
// set Global and Local work size dimensions
szGlobalWorkSize[0] = iTestN >> 3; // do 8 computations per work item
szLocalWorkSize[0]= iTestN>>3;
// Allocate and initialize host arrays
srcA = (void *)malloc (sizeof(cl_float) * iTestN);
srcB = (void *)malloc (sizeof(cl_float) * iTestN);
dst = (void *)malloc (sizeof(cl_float) * iTestN);
int i;
// Initialize arrays with some values
for (i=0;i<iTestN;i++)
{
((cl_float*)srcA)[i] = cl_float(i);
((cl_float*)srcB)[i] = 2;
((cl_float*)dst)[i]=-1;
}
cl_uint numPlatforms;
cl_platform_id platform = NULL;
cl_int status = clGetPlatformIDs(0, NULL, &numPlatforms);
if (0 < numPlatforms)
{
cl_platform_id* platforms = new cl_platform_id[numPlatforms];
status = clGetPlatformIDs(numPlatforms, platforms, NULL);
for (unsigned i = 0; i < numPlatforms; ++i)
{
char pbuf[100];
status = clGetPlatformInfo(platforms[i],
CL_PLATFORM_VENDOR,
sizeof(pbuf),
pbuf,
NULL);
platform = platforms[i];
if (!strcmp(pbuf, preferredPlatform))
{
printf("Found platform %s\n", preferredPlatform);
break;
}
}
delete[] platforms;
}
cl_context_properties cps[3] =
{
CL_CONTEXT_PLATFORM,
(cl_context_properties)platform,
0
};
// Create OpenCL context & context
cxGPUContext = clCreateContextFromType(cps, CL_DEVICE_TYPE_ALL, NULL, NULL, &ciErr1); //could also be CL_DEVICE_TYPE_GPU
// Query all devices available to the context
ciErr1 |= clGetContextInfo(cxGPUContext, CL_CONTEXT_DEVICES, 0, NULL, &szParmDataBytes);
cdDevices = (cl_device_id*)malloc(szParmDataBytes);
ciErr1 |= clGetContextInfo(cxGPUContext, CL_CONTEXT_DEVICES, szParmDataBytes, cdDevices, NULL);
if (cdDevices)
{
printDevInfo(cdDevices[0]);
}
// Create a command queue for first device the context reported
cqCommandQue = clCreateCommandQueue(cxGPUContext, cdDevices[0], 0, &ciErr2);
ciErr1 |= ciErr2;
// Allocate the OpenCL source and result buffer memory objects on the device GMEM
cmMemObjs[0] = clCreateBuffer(cxGPUContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float8) * szGlobalWorkSize[0], srcA, &ciErr2);
ciErr1 |= ciErr2;
cmMemObjs[1] = clCreateBuffer(cxGPUContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float8) * szGlobalWorkSize[0], srcB, &ciErr2);
ciErr1 |= ciErr2;
cmMemObjs[2] = clCreateBuffer(cxGPUContext, CL_MEM_WRITE_ONLY, sizeof(cl_float8) * szGlobalWorkSize[0], NULL, &ciErr2);
ciErr1 |= ciErr2;
///create kernels from binary
int numDevices = 1;
::size_t* lengths = (::size_t*) malloc(numDevices * sizeof(::size_t));
const unsigned char** images = (const unsigned char**) malloc(numDevices * sizeof(const void*));
for (i = 0; i < numDevices; ++i) {
images[i] = 0;
lengths[i] = 0;
}
// Read the OpenCL kernel in from source file
const char* cSourceFile = "opencl/vector_add/VectorAddKernels.cl";
const char* cPathAndName = cSourceFile;
#ifdef LOAD_FROM_FILE
size_t szKernelLength;
const char* cSourceCL =0;
char relativeFileName[1024];
{
const char* prefix[]={"../","../../","../../../","../../../../"};
int numPrefixes = sizeof(prefix)/sizeof(char*);
for (int i=0;!cSourceCL && i<numPrefixes;i++)
{
sprintf(relativeFileName,"%s%s",prefix[i],cSourceFile);
cSourceCL = loadProgSource(relativeFileName, "", &szKernelLength);
if (cSourceCL)
{
printf("Loaded program source: %s\n", relativeFileName);
}
}
}
if (!cSourceCL)
{
printf("Couldn't find file %s, exiting\n",cSourceFile);
exit(0);
}
#else
const char* cSourceCL = vectorAddCL;
size_t szKernelLength = strlen(cSourceCL);
#endif //LOAD_FROM_FILE
// Create the program
cpProgram = clCreateProgramWithSource(cxGPUContext, 1, (const char **)&cSourceCL, &szKernelLength, &ciErr1);
printf("clCreateProgramWithSource...\n");
if (ciErr1 != CL_SUCCESS)
{
printf("Error in clCreateProgramWithSource, Line %u in file %s !!!\n\n", __LINE__, __FILE__);
exit(0);
}
// Build the program with 'mad' Optimization option
#ifdef MAC
char* flags = "-cl-mad-enable -DMAC ";
#else
char flags[1024]={0};
#ifdef CL_PLATFORM_INTEL
sprintf(flags,"-g -s \"%s\"","C:/develop/experiments/opencl/vector_add/VectorAddKernels.cl");
#endif//CL_PLATFORM_INTEL
#endif//MAC
ciErr1 = clBuildProgram(cpProgram, 0, NULL, flags, NULL, NULL);
printf("clBuildProgram...\n");
if (ciErr1 != CL_SUCCESS)
{
printf("Error in clBuildProgram, Line %u in file %s !!!\n\n", __LINE__, __FILE__);
exit(0);
}
// Create the kernel
ckKernel = clCreateKernel(cpProgram, "VectorAdd", &ciErr1);
printf("clCreateKernel (VectorAdd)...\n");
if (ciErr1 != CL_SUCCESS)
{
printf("Error in clCreateKernel, Line %u in file %s !!!\n\n", __LINE__, __FILE__);
exit(0);
}
cl_int ciErrNum;
ciErrNum = clGetKernelWorkGroupInfo(ckKernel, cdDevices[0], CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &wgSize, NULL);
if (ciErrNum != CL_SUCCESS)
{
printf("cannot get workgroup size\n");
exit(0);
}
// Set the Argument values
ciErr1 |= clSetKernelArg(ckKernel, 0, sizeof(cl_mem), (void*)&cmMemObjs[0]);
ciErr1 |= clSetKernelArg(ckKernel, 1, sizeof(cl_mem), (void*)&cmMemObjs[1]);
ciErr1 |= clSetKernelArg(ckKernel, 2, sizeof(cl_mem), (void*)&cmMemObjs[2]);
ciErr1 |= clSetKernelArg(ckKernel, 3, sizeof(int), (void*)&actualGlobalSize);
printf("Press ENTER to quit\n");
getchar();
int workgroupSize = wgSize;
if(workgroupSize <= 0)
{ // let OpenCL library calculate workgroup size
size_t globalWorkSize[2];
globalWorkSize[0] = actualGlobalSize;
globalWorkSize[1] = 1;
// Copy input data from host to GPU and launch kernel
ciErr1 |= clEnqueueNDRangeKernel(cqCommandQue, ckKernel, 1, NULL, globalWorkSize, NULL, 0,0,0 );
}
else
{
size_t localWorkSize[2], globalWorkSize[2];
//workgroupSize = btMin(workgroupSize, actualGlobalSize);
int num_t = actualGlobalSize / workgroupSize;
int num_g = num_t * workgroupSize;
if(num_g < actualGlobalSize)
{
num_t++;
//this can cause problems -> processing outside of the buffer
//make sure to check kernel
}
size_t globalThreads[] = {num_t * workgroupSize};
size_t localThreads[] = {workgroupSize};
localWorkSize[0] = workgroupSize;
globalWorkSize[0] = num_t * workgroupSize;
localWorkSize[1] = 1;
globalWorkSize[1] = 1;
// Copy input data from host to GPU and launch kernel
ciErr1 |= clEnqueueNDRangeKernel(cqCommandQue, ckKernel, 1, NULL, globalThreads, localThreads, 0, NULL, NULL);
}
if (ciErrNum != CL_SUCCESS)
{
printf("cannot clEnqueueNDRangeKernel\n");
exit(0);
}
clFinish(cqCommandQue);
// Read back results and check accumulated errors
ciErr1 |= clEnqueueReadBuffer(cqCommandQue, cmMemObjs[2], CL_TRUE, 0, sizeof(cl_float8) * szGlobalWorkSize[0], dst, 0, NULL, NULL);
// Release kernel, program, and memory objects
// NOTE: Most properly this should be done at any of the exit points above, but it is omitted elsewhere for clarity.
free(cdDevices);
clReleaseKernel(ckKernel);
clReleaseProgram(cpProgram);
clReleaseCommandQueue(cqCommandQue);
clReleaseContext(cxGPUContext);
// print the results
int iErrorCount = 0;
for (i = 0; i < iTestN; i++)
{
if (((float*)dst)[i] != ((float*)srcA)[i]+((float*)srcB)[i])
iErrorCount++;
}
if (iErrorCount)
{
printf("Validation FAILED\n");
} else
{
printf("Validation SUCCESSFULL\n");
}
// Free host memory, close log and return success
for (i = 0; i < 3; i++)
{
clReleaseMemObject(cmMemObjs[i]);
}
free(srcA);
free(srcB);
free (dst);
printf("Press ENTER to quit\n");
getchar();
}

28
opencl/vector_add/premake4.lua
View File

@@ -1,28 +0,0 @@
function createProject(vendor)
hasCL = findOpenCL(vendor)
if (hasCL) then
project ("OpenCL_VectorAdd_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../bin"
files {
"main.cpp",
"../basic_initialize/b3OpenCLUtils.cpp",
"../basic_initialize/b3OpenCLUtils.h"
}
end
end
createProject("AMD")
createProject("Intel")
createProject("NVIDIA")
createProject("Apple")

69
opencl/vector_add_simplified/main.cpp
View File

@@ -1,69 +0,0 @@
///original author: Erwin Coumans
#include "b3OpenCLUtils.h"
#include "../parallel_primitives/host/btOpenCLArray.h"
#include "../parallel_primitives/host/btLauncherCL.h"
#include <stdio.h>
#define MSTRINGIFY(A) #A
const char* kernelString= MSTRINGIFY(
__kernel void VectorAdd(__global const float* a, __global const float* b, __global float* c, int numElements)
{
int iGID = get_global_id(0);
if (iGID>=numElements)
return;
float aGID = a[iGID];
float bGID = b[iGID];
float result = aGID + bGID;
c[iGID] = result;
}
);
int main(int argc, char* argv[])
{
int ciErrNum = 0;
int preferred_device = -1;
int preferred_platform = -1;
cl_platform_id platformId;
cl_context ctx;
cl_command_queue queue;
cl_device_id device;
cl_kernel addKernel;
ctx = b3OpenCLUtils::createContextFromType(CL_DEVICE_TYPE_GPU, &ciErrNum,0,0,preferred_device,preferred_platform,&platformId);
b3OpenCLUtils::printPlatformInfo(platformId);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
if (!ctx) {
printf("No OpenCL capable GPU found!");
return 0;
}
device = b3OpenCLUtils::getDevice(ctx,0);
queue = clCreateCommandQueue(ctx, device, 0, &ciErrNum);
addKernel = b3OpenCLUtils::compileCLKernelFromString(ctx,device,kernelString,"VectorAdd",&ciErrNum);
oclCHECKERROR(ciErrNum, CL_SUCCESS);
int numElements = 32;
btOpenCLArray<float> a(ctx,queue);
btOpenCLArray<float> b(ctx,queue);
btOpenCLArray<float> c(ctx,queue);
for (int i=0;i<numElements;i++)
{
a.push_back(float(i));
b.push_back(float(i));
}
c.resize(numElements);
btLauncherCL launcher( queue, addKernel);
launcher.setBuffer( a.getBufferCL());
launcher.setBuffer( b.getBufferCL());
launcher.setBuffer( c.getBufferCL());
launcher.setConst( numElements );
launcher.launch1D( numElements);
for (int i=0;i<numElements;i++)
{
float v = c.at(i);
printf("c[%d]=%f\n",i,v);
}
clReleaseCommandQueue(queue);
clReleaseContext(ctx);
return 0;
}

41
opencl/vector_add_simplified/premake4.lua
View File

@@ -1,41 +0,0 @@
function createProject (vendor)
local hasCL = findOpenCL(vendor)
if (hasCL) then
project ( "OpenCL_vector_add_simplified_" .. vendor)
initOpenCL(vendor)
language "C++"
kind "ConsoleApp"
targetdir "../../bin"
links {
"OpenCL_lib_parallel_primitives_host_" .. vendor
}
includedirs {
"../basic_initialize",
"../../src"
}
files {
"main.cpp",
"../basic_initialize/b3OpenCLUtils.cpp",
"../basic_initialize/b3OpenCLUtils.h",
"../../src/Bullet3Common/b3AlignedAllocator.cpp",
"../../src/Bullet3Common/b3AlignedAllocator.h",
"../../src/Bullet3Common/b3AlignedObjectArray.h",
}
end
end
createProject("AMD")
createProject("NVIDIA")
createProject("Intel")
createProject("Apple")

26
src/Bullet3Collision/BroadPhaseCollision/b3BroadphaseCallback.h
View File

@@ -1,26 +1,40 @@
/*
Bullet Continuous Collision Detection and Physics Library
Copyright (c) 2003-2013 Erwin Coumans 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.
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 B3_BROADPHASE_CALLBACK_H
#define B3_BROADPHASE_CALLBACK_H
#include "Bullet3Common/b3Vector3.h"
struct btBroadphaseProxy;
struct b3BroadphaseProxy;
struct btBroadphaseAabbCallback
struct b3BroadphaseAabbCallback
{
virtual ~btBroadphaseAabbCallback() {}
virtual bool process(const btBroadphaseProxy* proxy) = 0;
virtual ~b3BroadphaseAabbCallback() {}
virtual bool process(const b3BroadphaseProxy* proxy) = 0;
};
struct btBroadphaseRayCallback : public btBroadphaseAabbCallback
struct b3BroadphaseRayCallback : public b3BroadphaseAabbCallback
{
///added some cached data to accelerate ray-AABB tests
b3Vector3 m_rayDirectionInverse;
unsigned int m_signs[3];
b3Scalar m_lambda_max;
virtual ~btBroadphaseRayCallback() {}
virtual ~b3BroadphaseRayCallback() {}
};
#endif //B3_BROADPHASE_CALLBACK_H

436
src/Bullet3Collision/BroadPhaseCollision/b3DynamicBvh.cpp
View File

File diff suppressed because it is too large Load Diff

636
src/Bullet3Collision/BroadPhaseCollision/b3DynamicBvh.h
View File

File diff suppressed because it is too large Load Diff

Some files were not shown because too many files have changed in this diff Show More