AuroraMiddleware/OpenSubdiv
1
0
Fork 0
mirror of https://github.com/PixarAnimationStudios/OpenSubdiv synced 2024-11-27 14:00:10 +00:00
Code Issues Releases Wiki Activity

KernelBatch tables refactor / cleanup :

Browse Source 
- FarKernelBatch becomes a class w/ accessors
- split the FarKernelBatchFactory to its own header file
- add doxy doc
- propagate fallout to the rest of the code base
This commit is contained in:
manuelk 2013-05-06 17:50:58 -07:00
parent 7554321413
commit aed197628c
15 changed files with 794 additions and 393 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

1
opensubdiv/far/CMakeLists.txt
View File

@ -62,6 +62,7 @@ set(PUBLIC_HEADER_FILES
catmarkSubdivisionTablesFactory.h
dispatcher.h
kernelBatch.h
kernelBatchFactory.h
loopSubdivisionTables.h
loopSubdivisionTablesFactory.h
meshFactory.h

20
opensubdiv/far/bilinearSubdivisionTablesFactory.h
View File

@ -81,6 +81,14 @@ protected:
template <class X, class Y> friend class FarMeshFactory;
/// Creates a FarBilinearSubdivisiontables instance.
///
/// @param meshFactory a valid FarMeshFactory instance
///
/// @param farMesh
///
/// @param batches a vector of Kernel refinement batches : the factory
/// will reserve and append refinement tasks
///
static FarBilinearSubdivisionTables<U> * Create( FarMeshFactory<T,U> * meshFactory, FarMesh<U> * farMesh, FarKernelBatchVector *batches );
};
@ -134,8 +142,8 @@ FarBilinearSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFact
// "For each vertex, gather all the vertices from the parent face."
int nFaceVertices = (int)tablesFactory._faceVertsList[level].size();
if (nFaceVertices > 0)
batches->push_back(FarKernelBatch(level,
BILINEAR_FACE_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::BILINEAR_FACE_VERTEX,
level,
/*tableIndex=*/0,
/*start=*/0,
/*end=*/nFaceVertices,
@ -167,8 +175,8 @@ FarBilinearSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFact
// "Average the end-points of the parent edge"
int nEdgeVertices = (int)tablesFactory._edgeVertsList[level].size();
if (nEdgeVertices > 0)
batches->push_back(FarKernelBatch(level,
BILINEAR_EDGE_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::BILINEAR_EDGE_VERTEX,
level,
/*tableIndex=*/0,
/*start=*/0,
/*end=*/nEdgeVertices,
@ -194,8 +202,8 @@ FarBilinearSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFact
// "Pass down the parent vertex"
int nVertVertices = (int)tablesFactory._vertVertsList[level].size();
batches->push_back(FarKernelBatch(level,
BILINEAR_VERT_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::BILINEAR_VERT_VERTEX,
level,
/*tableIndex=*/0,
/*start=*/0,
/*end=*/nVertVertices,

19
opensubdiv/far/catmarkSubdivisionTablesFactory.h
View File

@ -65,6 +65,7 @@
#include "../far/catmarkSubdivisionTables.h"
#include "../far/meshFactory.h"
#include "../far/kernelBatchFactory.h"
#include "../far/subdivisionTablesFactory.h"
namespace OpenSubdiv {
@ -81,6 +82,14 @@ protected:
template <class X, class Y> friend class FarMeshFactory;
/// Creates a FarCatmarkSubdivisiontables instance.
///
/// @param meshFactory a valid FarMeshFactory instance
///
/// @param farMesh
///
/// @param batches a vector of Kernel refinement batches : the factory
/// will reserve and append refinement tasks
///
static FarCatmarkSubdivisionTables<U> * Create( FarMeshFactory<T,U> * meshFactory, FarMesh<U> * farMesh, FarKernelBatchVector *batches );
};
@ -144,8 +153,8 @@ FarCatmarkSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFacto
// add a batch for face vertices
if (nFaceVertices > 0) // in torus case, nfacevertices could be zero
batches->push_back(FarKernelBatch(level,
CATMARK_FACE_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::CATMARK_FACE_VERTEX,
level,
0,
0,
nFaceVertices,
@ -187,8 +196,8 @@ FarCatmarkSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFacto
// add a batch for edge vertices
if (nEdgeVertices > 0)
batches->push_back(FarKernelBatch(level,
CATMARK_EDGE_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::CATMARK_EDGE_VERTEX,
level,
0,
0,
nEdgeVertices,
@ -356,7 +365,7 @@ FarCatmarkSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFacto
// add batches for vert vertices
if (nVertVertices > 0)
batchFactory.AppendCatmarkBatches(batches, level, vertTableOffset, vertexOffset);
batchFactory.AppendCatmarkBatches(level, vertTableOffset, vertexOffset, batches);
vertexOffset += nVertVertices;
vertTableOffset += nVertVertices;
}

166
opensubdiv/far/dispatcher.h
View File

@ -97,49 +97,49 @@ FarDispatcher::Refine(CONTROLLER const *controller, FarKernelBatchVector const &
for (int i = 0; i < (int)batches.size(); ++i) {
const FarKernelBatch &batch = batches[i];
if (maxlevel >= 0 && batch.level >= maxlevel) continue;
if (maxlevel >= 0 && batch.GetLevel() >= maxlevel) continue;
switch(batch.kernelType) {
case CATMARK_FACE_VERTEX:
switch(batch.GetKernelType()) {
case FarKernelBatch::CATMARK_FACE_VERTEX:
controller->ApplyCatmarkFaceVerticesKernel(batch, clientdata);
break;
case CATMARK_EDGE_VERTEX:
case FarKernelBatch::CATMARK_EDGE_VERTEX:
controller->ApplyCatmarkEdgeVerticesKernel(batch, clientdata);
break;
case CATMARK_VERT_VERTEX_B:
case FarKernelBatch::CATMARK_VERT_VERTEX_B:
controller->ApplyCatmarkVertexVerticesKernelB(batch, clientdata);
break;
case CATMARK_VERT_VERTEX_A1:
case FarKernelBatch::CATMARK_VERT_VERTEX_A1:
controller->ApplyCatmarkVertexVerticesKernelA1(batch, clientdata);
break;
case CATMARK_VERT_VERTEX_A2:
case FarKernelBatch::CATMARK_VERT_VERTEX_A2:
controller->ApplyCatmarkVertexVerticesKernelA2(batch, clientdata);
break;
case LOOP_EDGE_VERTEX:
case FarKernelBatch::LOOP_EDGE_VERTEX:
controller->ApplyLoopEdgeVerticesKernel(batch, clientdata);
break;
case LOOP_VERT_VERTEX_B:
case FarKernelBatch::LOOP_VERT_VERTEX_B:
controller->ApplyLoopVertexVerticesKernelB(batch, clientdata);
break;
case LOOP_VERT_VERTEX_A1:
case FarKernelBatch::LOOP_VERT_VERTEX_A1:
controller->ApplyLoopVertexVerticesKernelA1(batch, clientdata);
break;
case LOOP_VERT_VERTEX_A2:
case FarKernelBatch::LOOP_VERT_VERTEX_A2:
controller->ApplyLoopVertexVerticesKernelA2(batch, clientdata);
break;
case BILINEAR_FACE_VERTEX:
case FarKernelBatch::BILINEAR_FACE_VERTEX:
controller->ApplyBilinearFaceVerticesKernel(batch, clientdata);
break;
case BILINEAR_EDGE_VERTEX:
case FarKernelBatch::BILINEAR_EDGE_VERTEX:
controller->ApplyBilinearEdgeVerticesKernel(batch, clientdata);
break;
case BILINEAR_VERT_VERTEX:
case FarKernelBatch::BILINEAR_VERT_VERTEX:
controller->ApplyBilinearVertexVerticesKernel(batch, clientdata);
break;
case HIERARCHICAL_EDIT:
case FarKernelBatch::HIERARCHICAL_EDIT:
controller->ApplyVertexEdits(batch, clientdata);
break;
}
@ -152,6 +152,7 @@ FarDispatcher::Refine(CONTROLLER const *controller, FarKernelBatchVector const &
template <class U>
class FarComputeController
{
public:
void Refine(FarMesh<U> * mesh, int maxlevel=-1) const;
@ -183,6 +184,9 @@ public:
void ApplyVertexEdits(FarKernelBatch const &batch, void * clientdata) const;
static FarComputeController _DefaultController;
private:
};
template<class U> FarComputeController<U> FarComputeController<U>::_DefaultController;
@ -195,118 +199,224 @@ FarComputeController<U>::Refine(FarMesh<U> *mesh, int maxlevel) const {
template <class U> void
FarComputeController<U>::ApplyBilinearFaceVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarBilinearSubdivisionTables<U> const * subdivision =
dynamic_cast<FarBilinearSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeFacePoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeFacePoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyBilinearEdgeVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarBilinearSubdivisionTables<U> const * subdivision =
dynamic_cast<FarBilinearSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeEdgePoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeEdgePoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyBilinearVertexVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarBilinearSubdivisionTables<U> const * subdivision =
dynamic_cast<FarBilinearSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyCatmarkFaceVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarCatmarkSubdivisionTables<U> const * subdivision =
dynamic_cast<FarCatmarkSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeFacePoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeFacePoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyCatmarkEdgeVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarCatmarkSubdivisionTables<U> const * subdivision =
dynamic_cast<FarCatmarkSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeEdgePoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeEdgePoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyCatmarkVertexVerticesKernelB(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarCatmarkSubdivisionTables<U> const * subdivision =
dynamic_cast<FarCatmarkSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsB(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsB( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyCatmarkVertexVerticesKernelA1(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarCatmarkSubdivisionTables<U> const * subdivision =
dynamic_cast<FarCatmarkSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsA(batch.vertexOffset, false, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsA( batch.GetVertexOffset(),
false,
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyCatmarkVertexVerticesKernelA2(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarCatmarkSubdivisionTables<U> const * subdivision =
dynamic_cast<FarCatmarkSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsA(batch.vertexOffset, true, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsA( batch.GetVertexOffset(),
true,
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyLoopEdgeVerticesKernel(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarLoopSubdivisionTables<U> const * subdivision =
dynamic_cast<FarLoopSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeEdgePoints(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeEdgePoints( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyLoopVertexVerticesKernelB(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarLoopSubdivisionTables<U> const * subdivision =
dynamic_cast<FarLoopSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsB(batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsB( batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyLoopVertexVerticesKernelA1(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarLoopSubdivisionTables<U> const * subdivision =
dynamic_cast<FarLoopSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsA(batch.vertexOffset, false, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsA( batch.GetVertexOffset(),
false,
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
template <class U> void
FarComputeController<U>::ApplyLoopVertexVerticesKernelA2(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarLoopSubdivisionTables<U> const * subdivision =
dynamic_cast<FarLoopSubdivisionTables<U> const *>(mesh->GetSubdivisionTables());
assert(subdivision);
subdivision->computeVertexPointsA(batch.vertexOffset, true, batch.tableOffset, batch.start, batch.end, clientdata);
subdivision->computeVertexPointsA( batch.GetVertexOffset(),
true,
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata);
}
template <class U> void
FarComputeController<U>::ApplyVertexEdits(FarKernelBatch const &batch, void * clientdata) const {
FarMesh<U> * mesh = static_cast<FarMesh<U> *>(clientdata);
FarVertexEditTables<U> const * vertEdit = mesh->GetVertexEdit();
if (vertEdit)
vertEdit->computeVertexEdits(batch.tableIndex, batch.vertexOffset, batch.tableOffset, batch.start, batch.end, clientdata);
vertEdit->computeVertexEdits( batch.GetTableIndex(),
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
clientdata );
}
} // end namespace OPENSUBDIV_VERSION

243
opensubdiv/far/kernelBatch.h
View File

@ -64,6 +64,38 @@
namespace OpenSubdiv {
namespace OPENSUBDIV_VERSION {
/// \brief A GP Compute Kernel descriptor.
///
/// Vertex refinement through subdivision schemes requires the successive
/// application of dedicated compute kernels. OpenSubdiv groups these vertices
/// in batches based on their topology in order to minimize the number of kernel
/// switches to process a given primitive.
///
/// [Subdivision table for kernel k]
/// ------+---------------------------------+-----
/// | Prim p, Level n |
/// ------+---------------------------------+-----
/// | |<batch range>| |
/// ------+---------------------------------+-----
/// ^ ^ ^
/// tableOffset start end
/// . .
/// . .
/// . .
/// [VBO] . .
/// ------+---------------------------------+-----
/// | Prim p, Kernel k, Level n |
/// ------+---------------------------------+-----
/// | |<batch range>| |
/// ------+---------------------------------+-----
/// ^ ^ ^
/// vertexOffset start end
///
class FarKernelBatch {
public:
enum KernelType {
CATMARK_FACE_VERTEX,
CATMARK_EDGE_VERTEX,
@ -80,120 +112,119 @@ enum KernelType {
HIERARCHICAL_EDIT,
};
struct FarKernelBatch {
FarKernelBatch(int level_,
int kernelType_,
int tableIndex_,
int start_,
int end_,
int tableOffset_,
int vertexOffset_) :
level(level_),
kernelType(kernelType_),
tableIndex(tableIndex_),
start(start_),
end(end_),
tableOffset(tableOffset_),
vertexOffset(vertexOffset_) {
/// Constructor.
///
/// @param kernelType the type of compute kernel kernel
///
/// @param level the level of subdivision of the vertices in the batch
///
/// @param tableIndex edit index (for the hierarchical edit kernels only)
///
/// @param start index of the first vertex in the batch
///
/// @param end index of the last vertex in the batch
///
/// @param tableOffset XXXX
///
/// @param vertexOffset XXXX
///
FarKernelBatch( KernelType kernelType,
int level,
int tableIndex,
int start,
int end,
int tableOffset,
int vertexOffset) :
_kernelType(kernelType),
_level(level),
_tableIndex(tableIndex),
_start(start),
_end(end),
_tableOffset(tableOffset),
_vertexOffset(vertexOffset) {
}
/*
[Subdivision table for kernel k]
------+---------------------------------+-----
| Prim p, Level n |
------+---------------------------------+-----
| |<batch range>| |
------+---------------------------------+-----
^ ^ ^
tableOffset start end
. .
. .
. .
[VBO] . .
------+---------------------------------+-----
| Prim p, Kernel k, Level n |
------+---------------------------------+-----
| |<batch range>| |
------+---------------------------------+-----
^ ^ ^
vertexOffset start end
*/
/// Returns the type of kernel to apply to the vertices in the batch.
KernelType GetKernelType() const {
return _kernelType;
}
int level;
int kernelType;
int tableIndex; // edit index (for the h-edit kernel only)
int start;
int end;
int tableOffset;
int vertexOffset;
/// Returns the subdivision level of the vertices in the batch
int GetLevel() const {
return _level;
}
/// Returns the index of the first vertex in the batch
int GetStart() const {
return _start;
}
/// Returns the index of the first vertex in the batch
const int * GetStartPtr() const {
return & _start;
}
/// Returns the index of the last vertex in the batch
int GetEnd() const {
return _end;
}
/// Returns the index of the last vertex in the batch
const int * GetEndPtr() const {
return & _end;
}
/// Returns the edit index (for the hierarchical edit kernels only)
int GetTableIndex() const {
return _tableIndex;
}
/// Returns
int GetTableOffset() const {
return _tableOffset;
}
/// Returns
const int * GetTableOffsetPtr() const {
return & _tableOffset;
}
/// Returns
int GetVertexOffset() const {
return _vertexOffset;
}
/// Returns
const int * GetVertexOffsetPtr() const {
return & _vertexOffset;
}
private:
friend class FarKernelBatchFactory;
template <class X, class Y> friend class FarMultiMeshFactory;
KernelType _kernelType;
int _level;
int _tableIndex;
int _start;
int _end;
int _tableOffset;
int _vertexOffset;
};
typedef std::vector<FarKernelBatch> FarKernelBatchVector;
struct FarVertexKernelBatchFactory {
FarVertexKernelBatchFactory(int a, int b) {
kernelB.first = kernelA1.first = kernelA2.first = a;
kernelB.second = kernelA1.second = kernelA2.second = b;
}
void AddVertex( int index, int rank ) {
// expand the range of kernel batches based on vertex index and rank
if (rank<7) {
if (index < kernelB.first)
kernelB.first=index;
if (index > kernelB.second)
kernelB.second=index;
}
if ((rank>2) and (rank<8)) {
if (index < kernelA2.first)
kernelA2.first=index;
if (index > kernelA2.second)
kernelA2.second=index;
}
if (rank>6) {
if (index < kernelA1.first)
kernelA1.first=index;
if (index > kernelA1.second)
kernelA1.second=index;
}
}
void AppendCatmarkBatches(FarKernelBatchVector *result, int level, int tableOffset, int vertexOffset) {
if (kernelB.second >= kernelB.first)
result->push_back(FarKernelBatch(level, CATMARK_VERT_VERTEX_B, 0,
kernelB.first, kernelB.second+1,
tableOffset, vertexOffset));
if (kernelA1.second >= kernelA1.first)
result->push_back(FarKernelBatch(level, CATMARK_VERT_VERTEX_A1, 0,
kernelA1.first, kernelA1.second+1,
tableOffset, vertexOffset));
if (kernelA2.second >= kernelA2.first)
result->push_back(FarKernelBatch(level, CATMARK_VERT_VERTEX_A2, 0,
kernelA2.first, kernelA2.second+1,
tableOffset, vertexOffset));
}
void AppendLoopBatches(FarKernelBatchVector *result, int level, int tableOffset, int vertexOffset) {
if (kernelB.second >= kernelB.first)
result->push_back(FarKernelBatch(level, LOOP_VERT_VERTEX_B, 0,
kernelB.first, kernelB.second+1,
tableOffset, vertexOffset));
if (kernelA1.second >= kernelA1.first)
result->push_back(FarKernelBatch(level, LOOP_VERT_VERTEX_A1, 0,
kernelA1.first, kernelA1.second+1,
tableOffset, vertexOffset));
if (kernelA2.second >= kernelA2.first)
result->push_back(FarKernelBatch(level, LOOP_VERT_VERTEX_A2, 0,
kernelA2.first, kernelA2.second+1,
tableOffset, vertexOffset));
}
std::pair<int,int> kernelB; // first / last vertex vertex batch (kernel B)
std::pair<int,int> kernelA1; // first / last vertex vertex batch (kernel A pass 1)
std::pair<int,int> kernelA2; // first / last vertex vertex batch (kernel A pass 2)
};
} // end namespace OPENSUBDIV_VERSION
using namespace OPENSUBDIV_VERSION;

219
opensubdiv/far/kernelBatchFactory.h Normal file
View File

@ -0,0 +1,219 @@
//
// Copyright (C) Pixar. All rights reserved.
//
// This license governs use of the accompanying software. If you
// use the software, you accept this license. If you do not accept
// the license, do not use the software.
//
// 1. Definitions
// The terms "reproduce," "reproduction," "derivative works," and
// "distribution" have the same meaning here as under U.S.
// copyright law. A "contribution" is the original software, or
// any additions or changes to the software.
// A "contributor" is any person or entity that distributes its
// contribution under this license.
// "Licensed patents" are a contributor's patent claims that read
// directly on its contribution.
//
// 2. Grant of Rights
// (A) Copyright Grant- Subject to the terms of this license,
// including the license conditions and limitations in section 3,
// each contributor grants you a non-exclusive, worldwide,
// royalty-free copyright license to reproduce its contribution,
// prepare derivative works of its contribution, and distribute
// its contribution or any derivative works that you create.
// (B) Patent Grant- Subject to the terms of this license,
// including the license conditions and limitations in section 3,
// each contributor grants you a non-exclusive, worldwide,
// royalty-free license under its licensed patents to make, have
// made, use, sell, offer for sale, import, and/or otherwise
// dispose of its contribution in the software or derivative works
// of the contribution in the software.
//
// 3. Conditions and Limitations
// (A) No Trademark License- This license does not grant you
// rights to use any contributor's name, logo, or trademarks.
// (B) If you bring a patent claim against any contributor over
// patents that you claim are infringed by the software, your
// patent license from such contributor to the software ends
// automatically.
// (C) If you distribute any portion of the software, you must
// retain all copyright, patent, trademark, and attribution
// notices that are present in the software.
// (D) If you distribute any portion of the software in source
// code form, you may do so only under this license by including a
// complete copy of this license with your distribution. If you
// distribute any portion of the software in compiled or object
// code form, you may only do so under a license that complies
// with this license.
// (E) The software is licensed "as-is." You bear the risk of
// using it. The contributors give no express warranties,
// guarantees or conditions. You may have additional consumer
// rights under your local laws which this license cannot change.
// To the extent permitted under your local laws, the contributors
// exclude the implied warranties of merchantability, fitness for
// a particular purpose and non-infringement.
//
#ifndef FAR_KERNEL_BATCH_FACTORY_H
#define FAR_KERNEL_BATCH_FACTORY_H
#include "../version.h"
#include "../far/kernelBatch.h"
namespace OpenSubdiv {
namespace OPENSUBDIV_VERSION {
class FarVertexKernelBatchFactory {
public:
/// Constructor.
///
/// @param start index of the first vertex in the batch
///
/// @param end index of the last vertex in the batch
///
FarVertexKernelBatchFactory(int start, int end) {
kernelB.start = kernelA1.start = kernelA2.start = start;
kernelB.end = kernelA1.end = kernelA2.end = end;
}
/// Adds a vertex-vertex to the appropriate compute batch based on "Rank".
/// Ranking is based on the interpolation required (Smooth, Dart, Crease,
/// or Corner). With semi-sharp creases, two passes of interpolation are
/// required, from which we derive a matrix of compute kernel combinations.
///
/// The kernel combinatorial matrix :
///
/// Rules +----+----+----+----+----+----+----+----+----+----+
/// Pass 0 | Dt | Sm | Sm | Dt | Sm | Dt | Sm | Cr | Co | Cr |
/// Pass 1 | | | | Co | Co | Cr | Cr | Co | | |
/// Kernel +----+----+----+----+----+----+----+----+----+----+
/// Pass 0 | B | B | B | B | B | B | B | A | A | A |
/// Pass 1 | | | | A | A | A | A | A | | |
/// +----+----+----+----+----+----+----+----+----+----+
/// Rank | 0 | 1 | 2 | 3 | 4 | 5 | 6 | 7 | 8 | 9 |
/// +----+----+----+----+----+----+----+----+----+----+
/// with :
/// - A : compute kernel applying k_Crease / k_Corner rules
/// - B : compute kernel applying k_Smooth / k_Dart rules
///
/// @param index the index of the vertex
///
/// @param rank the rank of the vertex (see
/// FarSubdivisionTables::GetMaskRanking())
///
void AddVertex( int index, int rank );
/// Appends a FarKernelBatch to a vector of batches for Catmark subdivision
///
/// @param level the subdivision level of the vertices in the batch
///
/// @param tableOffset XXXX
///
/// @param vertexOffset XXXX
///
/// @result the expanded batch vector
///
void AppendCatmarkBatches(int level, int tableOffset, int vertexOffset, FarKernelBatchVector *result);
/// Appends a FarKernelBatch to a vector of batches for Loop subdivision
///
/// @param level the subdivision level of the vertices in the batch
///
/// @param tableOffset XXXX
///
/// @param vertexOffset XXXX
///
/// @result the expanded batch vector
///
void AppendLoopBatches(int level, int tableOffset, int vertexOffset, FarKernelBatchVector *result);
private:
struct Range {
int start,
end;
};
Range kernelB; // vertex batch reange (kernel B)
Range kernelA1; // vertex batch reange (kernel A pass 1)
Range kernelA2; // vertex batch reange (kernel A pass 2)
};
inline void
FarVertexKernelBatchFactory::AddVertex( int index, int rank ) {
// expand the range of kernel batches based on vertex index and rank
if (rank<7) {
if (index < kernelB.start)
kernelB.start=index;
if (index > kernelB.end)
kernelB.end=index;
}
if ((rank>2) and (rank<8)) {
if (index < kernelA2.start)
kernelA2.start=index;
if (index > kernelA2.end)
kernelA2.end=index;
}
if (rank>6) {
if (index < kernelA1.start)
kernelA1.start=index;
if (index > kernelA1.end)
kernelA1.end=index;
}
}
inline void
FarVertexKernelBatchFactory::AppendCatmarkBatches(int level,
int tableOffset,
int vertexOffset,
FarKernelBatchVector *result) {
if (kernelB.end >= kernelB.start)
result->push_back(FarKernelBatch( FarKernelBatch::CATMARK_VERT_VERTEX_B, level, 0,
kernelB.start, kernelB.end+1,
tableOffset, vertexOffset) );
if (kernelA1.end >= kernelA1.start)
result->push_back(FarKernelBatch( FarKernelBatch::CATMARK_VERT_VERTEX_A1, level, 0,
kernelA1.start, kernelA1.end+1,
tableOffset, vertexOffset));
if (kernelA2.end >= kernelA2.start)
result->push_back(FarKernelBatch( FarKernelBatch::CATMARK_VERT_VERTEX_A2, level, 0,
kernelA2.start, kernelA2.end+1,
tableOffset, vertexOffset) );
}
inline void
FarVertexKernelBatchFactory::AppendLoopBatches(int level,
int tableOffset,
int vertexOffset,
FarKernelBatchVector *result) {
if (kernelB.end >= kernelB.start)
result->push_back(FarKernelBatch( FarKernelBatch::LOOP_VERT_VERTEX_B, level, 0,
kernelB.start, kernelB.end+1,
tableOffset, vertexOffset) );
if (kernelA1.end >= kernelA1.start)
result->push_back(FarKernelBatch( FarKernelBatch::LOOP_VERT_VERTEX_A1, level, 0,
kernelA1.start, kernelA1.end+1,
tableOffset, vertexOffset) );
if (kernelA2.end >= kernelA2.start)
result->push_back(FarKernelBatch( FarKernelBatch::LOOP_VERT_VERTEX_A2, level, 0,
kernelA2.start, kernelA2.end+1,
tableOffset, vertexOffset) );
}
} // end namespace OPENSUBDIV_VERSION
using namespace OPENSUBDIV_VERSION;
} // end namespace OpenSubdiv
#endif /* FAR_KERNEL_BATCH_FACTORY_H */

15
opensubdiv/far/loopSubdivisionTablesFactory.h
View File

@ -62,6 +62,7 @@
#include "../far/loopSubdivisionTables.h"
#include "../far/meshFactory.h"
#include "../far/kernelBatchFactory.h"
#include "../far/subdivisionTablesFactory.h"
#include <cassert>
@ -81,6 +82,14 @@ protected:
template <class X, class Y> friend class FarMeshFactory;
/// Creates a FarLoopSubdivisiontables instance.
///
/// @param meshFactory a valid FarMeshFactory instance
///
/// @param farMesh
///
/// @param batches a vector of Kernel refinement batches : the factory
/// will reserve and append refinement tasks
///
static FarLoopSubdivisionTables<U> * Create( FarMeshFactory<T,U> * meshFactory, FarMesh<U> * farMesh, FarKernelBatchVector * batches );
};
@ -134,8 +143,8 @@ FarLoopSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFactory,
// Edge vertices
int nEdgeVertices = (int)tablesFactory._edgeVertsList[level].size();
if (nEdgeVertices > 0)
batches->push_back(FarKernelBatch(level,
LOOP_EDGE_VERTEX,
batches->push_back(FarKernelBatch( FarKernelBatch::LOOP_EDGE_VERTEX,
level,
0,
0,
nEdgeVertices,
@ -286,7 +295,7 @@ FarLoopSubdivisionTablesFactory<T,U>::Create( FarMeshFactory<T,U> * meshFactory,
V_ITa += nVertVertices*5;
V_W += nVertVertices;
batchFactory.AppendLoopBatches(batches, level, vertTableOffset, vertexOffset);
batchFactory.AppendLoopBatches(level, vertTableOffset, vertexOffset, batches);
vertexOffset += nVertVertices;
vertTableOffset += nVertVertices;
}

43
opensubdiv/far/multiMeshFactory.h
View File

@ -332,25 +332,32 @@ FarMultiMeshFactory<T, U>::spliceSubdivisionTables(FarMesh<U> *farMesh, FarMeshV
for (size_t i = 0; i < meshes.size(); ++i) {
for (int j = 0; j < (int)meshes[i]->_batches.size(); ++j) {
FarKernelBatch batch = meshes[i]->_batches[j];
batch.vertexOffset += vertexOffsets[i];
batch._vertexOffset += vertexOffsets[i];
if (batch.kernelType == CATMARK_FACE_VERTEX or
batch.kernelType == BILINEAR_FACE_VERTEX) {
batch.tableOffset += fvOffsets[i];
} else if (batch.kernelType == CATMARK_EDGE_VERTEX or
batch.kernelType == LOOP_EDGE_VERTEX or
batch.kernelType == BILINEAR_EDGE_VERTEX) {
batch.tableOffset += evOffsets[i];
} else if (batch.kernelType == CATMARK_VERT_VERTEX_A1 or
batch.kernelType == CATMARK_VERT_VERTEX_A2 or
batch.kernelType == CATMARK_VERT_VERTEX_B or
batch.kernelType == LOOP_VERT_VERTEX_A1 or
batch.kernelType == LOOP_VERT_VERTEX_A2 or
batch.kernelType == LOOP_VERT_VERTEX_B or
batch.kernelType == BILINEAR_VERT_VERTEX) {
batch.tableOffset += vvOffsets[i];
} else if (batch.kernelType == HIERARCHICAL_EDIT) {
batch.tableIndex += editTableIndexOffset;
if (batch._kernelType == FarKernelBatch::CATMARK_FACE_VERTEX or
batch._kernelType == FarKernelBatch::BILINEAR_FACE_VERTEX) {
batch._tableOffset += fvOffsets[i];
} else if (batch._kernelType == FarKernelBatch::CATMARK_EDGE_VERTEX or
batch._kernelType == FarKernelBatch::LOOP_EDGE_VERTEX or
batch._kernelType == FarKernelBatch::BILINEAR_EDGE_VERTEX) {
batch._tableOffset += evOffsets[i];
} else if (batch._kernelType == FarKernelBatch::CATMARK_VERT_VERTEX_A1 or
batch._kernelType == FarKernelBatch::CATMARK_VERT_VERTEX_A2 or
batch._kernelType == FarKernelBatch::CATMARK_VERT_VERTEX_B or
batch._kernelType == FarKernelBatch::LOOP_VERT_VERTEX_A1 or
batch._kernelType == FarKernelBatch::LOOP_VERT_VERTEX_A2 or
batch._kernelType == FarKernelBatch::LOOP_VERT_VERTEX_B or
batch._kernelType == FarKernelBatch::BILINEAR_VERT_VERTEX) {
batch._tableOffset += vvOffsets[i];
} else if (batch._kernelType == FarKernelBatch::HIERARCHICAL_EDIT) {
batch._tableIndex += editTableIndexOffset;
}
batches.push_back(batch);
}

8
opensubdiv/far/vertexEditTablesFactory.h
View File

@ -82,6 +82,7 @@ protected:
/// Compares the number of subfaces in an edit (for sorting purposes)
static bool compareEdits(HbrVertexEdit<T> const *a, HbrVertexEdit<T> const *b);
static void insertHEditBatch(FarKernelBatchVector *batches, int batchIndex, int batchLevel, int batchCount, int tableOffset);
/// Creates a FarVertexEditTables instance.
@ -98,11 +99,14 @@ template <class T, class U> void
FarVertexEditTablesFactory<T,U>::insertHEditBatch(FarKernelBatchVector *batches, int batchIndex, int batchLevel, int batchCount, int tableOffset) {
FarKernelBatchVector::iterator it = batches->begin();
while (it != batches->end()) {
if (it->level > batchLevel+1) break;
if (it->GetLevel() > batchLevel+1)
break;
++it;
}
batches->insert(it, FarKernelBatch(batchLevel+1, HIERARCHICAL_EDIT, batchIndex, 0, batchCount, tableOffset, 0));
batches->insert(it, FarKernelBatch( FarKernelBatch::HIERARCHICAL_EDIT, batchLevel+1, batchIndex, 0, batchCount, tableOffset, 0) );
}
template <class T, class U> FarVertexEditTables<U> *

124
opensubdiv/osd/clComputeController.cpp
View File

@ -140,7 +140,7 @@ OsdCLComputeController::ApplyBilinearEdgeVerticesKernel(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetBilinearEdgeKernel();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -150,10 +150,10 @@ OsdCLComputeController::ApplyBilinearEdgeVerticesKernel(
clSetKernelArg(kernel, 0, sizeof(cl_mem), &vertexBuffer);
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &E_IT);
clSetKernelArg(kernel, 3, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 4, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.start);
clSetKernelArg(kernel, 6, sizeof(int), &batch.end);
clSetKernelArg(kernel, 3, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 4, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
NULL, 0, NULL, NULL);
@ -169,7 +169,7 @@ OsdCLComputeController::ApplyBilinearVertexVerticesKernel(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetBilinearVertexKernel();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -179,10 +179,10 @@ OsdCLComputeController::ApplyBilinearVertexVerticesKernel(
clSetKernelArg(kernel, 0, sizeof(cl_mem), &vertexBuffer);
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 4, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.start);
clSetKernelArg(kernel, 6, sizeof(int), &batch.end);
clSetKernelArg(kernel, 3, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 4, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
NULL, 0, NULL, NULL);
@ -198,7 +198,7 @@ OsdCLComputeController::ApplyCatmarkFaceVerticesKernel(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetCatmarkFaceKernel();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -210,10 +210,10 @@ OsdCLComputeController::ApplyCatmarkFaceVerticesKernel(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &F_IT);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &F_ITa);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
@ -230,7 +230,7 @@ OsdCLComputeController::ApplyCatmarkEdgeVerticesKernel(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetCatmarkEdgeKernel();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -242,10 +242,10 @@ OsdCLComputeController::ApplyCatmarkEdgeVerticesKernel(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &E_IT);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &E_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
@ -262,7 +262,7 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelB(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetCatmarkVertexKernelB();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -276,10 +276,10 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelB(
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_IT);
clSetKernelArg(kernel, 4, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 5, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 7, sizeof(int), &batch.start);
clSetKernelArg(kernel, 8, sizeof(int), &batch.end);
clSetKernelArg(kernel, 5, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 8, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
@ -296,7 +296,7 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelA1(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
int ipass = false;
cl_kernel kernel = context->GetKernelBundle()->GetCatmarkVertexKernelA();
@ -309,10 +309,10 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelA1(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
clSetKernelArg(kernel, 8, sizeof(int), &ipass);
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
@ -330,7 +330,7 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelA2(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
int ipass = true;
cl_kernel kernel = context->GetKernelBundle()->GetCatmarkVertexKernelA();
@ -343,10 +343,10 @@ OsdCLComputeController::ApplyCatmarkVertexVerticesKernelA2(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
clSetKernelArg(kernel, 8, sizeof(int), &ipass);
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
@ -364,7 +364,7 @@ OsdCLComputeController::ApplyLoopEdgeVerticesKernel(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetLoopEdgeKernel();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -376,10 +376,10 @@ OsdCLComputeController::ApplyLoopEdgeVerticesKernel(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &E_IT);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &E_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
@ -396,7 +396,7 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelB(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_kernel kernel = context->GetKernelBundle()->GetLoopVertexKernelB();
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
@ -410,10 +410,10 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelB(
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_IT);
clSetKernelArg(kernel, 4, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 5, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 7, sizeof(int), &batch.start);
clSetKernelArg(kernel, 8, sizeof(int), &batch.end);
clSetKernelArg(kernel, 5, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 8, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
@ -430,7 +430,7 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelA1(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
int ipass = false;
cl_kernel kernel = context->GetKernelBundle()->GetLoopVertexKernelA();
@ -443,10 +443,10 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelA1(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
clSetKernelArg(kernel, 8, sizeof(int), &ipass);
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
@ -464,7 +464,7 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelA2(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
int ipass = true;
cl_kernel kernel = context->GetKernelBundle()->GetLoopVertexKernelA();
@ -477,10 +477,10 @@ OsdCLComputeController::ApplyLoopVertexVerticesKernelA2(
clSetKernelArg(kernel, 1, sizeof(cl_mem), &varyingBuffer);
clSetKernelArg(kernel, 2, sizeof(cl_mem), &V_ITa);
clSetKernelArg(kernel, 3, sizeof(cl_mem), &V_W);
clSetKernelArg(kernel, 4, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 5, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.start);
clSetKernelArg(kernel, 7, sizeof(int), &batch.end);
clSetKernelArg(kernel, 4, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 5, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetEndPtr());
clSetKernelArg(kernel, 8, sizeof(int), &ipass);
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
@ -498,10 +498,10 @@ OsdCLComputeController::ApplyVertexEdits(
assert(context);
cl_int ciErrNum;
size_t globalWorkSize[1] = { batch.end - batch.start };
size_t globalWorkSize[1] = { batch.GetEnd() - batch.GetStart() };
cl_mem vertexBuffer = context->GetCurrentVertexBuffer();
const OsdCLHEditTable * edit = context->GetEditTable(batch.tableIndex);
const OsdCLHEditTable * edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
const OsdCLTable * primvarIndices = edit->GetPrimvarIndices();
@ -520,16 +520,16 @@ OsdCLComputeController::ApplyVertexEdits(
clSetKernelArg(kernel, 2, sizeof(cl_mem), &values);
clSetKernelArg(kernel, 3, sizeof(int), &primvarOffset);
clSetKernelArg(kernel, 4, sizeof(int), &primvarWidth);
clSetKernelArg(kernel, 5, sizeof(int), &batch.vertexOffset);
clSetKernelArg(kernel, 6, sizeof(int), &batch.tableOffset);
clSetKernelArg(kernel, 7, sizeof(int), &batch.start);
clSetKernelArg(kernel, 8, sizeof(int), &batch.end);
clSetKernelArg(kernel, 5, sizeof(int), batch.GetVertexOffsetPtr());
clSetKernelArg(kernel, 6, sizeof(int), batch.GetTableOffsetPtr());
clSetKernelArg(kernel, 7, sizeof(int), batch.GetStartPtr());
clSetKernelArg(kernel, 8, sizeof(int), batch.GetEndPtr());
ciErrNum = clEnqueueNDRangeKernel(context->GetCommandQueue(),
kernel, 1, NULL, globalWorkSize,
NULL, 0, NULL, NULL);
CL_CHECK_ERROR(ciErrNum, "vertex edit %d %d\n", batch.tableIndex, ciErrNum);
CL_CHECK_ERROR(ciErrNum, "vertex edit %d %d\n", batch.GetTableIndex(), ciErrNum);
} else if (edit->GetOperation() == FarVertexEdit::Set) {
// XXXX TODO

42
opensubdiv/osd/cpuComputeController.cpp
View File

@ -84,7 +84,7 @@ OsdCpuComputeController::ApplyBilinearFaceVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::F_IT)->GetBuffer(),
(const int*)context->GetTable(Table::F_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -100,7 +100,7 @@ OsdCpuComputeController::ApplyBilinearEdgeVerticesKernel(
context->GetCurrentVertexBuffer(),
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -116,7 +116,7 @@ OsdCpuComputeController::ApplyBilinearVertexVerticesKernel(
context->GetCurrentVertexBuffer(),
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -133,7 +133,7 @@ OsdCpuComputeController::ApplyCatmarkFaceVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::F_IT)->GetBuffer(),
(const int*)context->GetTable(Table::F_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -150,7 +150,7 @@ OsdCpuComputeController::ApplyCatmarkEdgeVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
(const float*)context->GetTable(Table::E_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -168,7 +168,7 @@ OsdCpuComputeController::ApplyCatmarkVertexVerticesKernelB(
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const int*)context->GetTable(Table::V_IT)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -185,7 +185,7 @@ OsdCpuComputeController::ApplyCatmarkVertexVerticesKernelA1(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -202,7 +202,7 @@ OsdCpuComputeController::ApplyCatmarkVertexVerticesKernelA2(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -219,7 +219,7 @@ OsdCpuComputeController::ApplyLoopEdgeVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
(const float*)context->GetTable(Table::E_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -237,7 +237,7 @@ OsdCpuComputeController::ApplyLoopVertexVerticesKernelB(
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const int*)context->GetTable(Table::V_IT)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -254,7 +254,7 @@ OsdCpuComputeController::ApplyLoopVertexVerticesKernelA1(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -271,7 +271,7 @@ OsdCpuComputeController::ApplyLoopVertexVerticesKernelA2(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -282,7 +282,7 @@ OsdCpuComputeController::ApplyVertexEdits(
static_cast<OsdCpuComputeContext*>(clientdata);
assert(context);
const OsdCpuHEditTable *edit = context->GetEditTable(batch.tableIndex);
const OsdCpuHEditTable *edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
const OsdCpuTable * primvarIndices = edit->GetPrimvarIndices();
@ -293,10 +293,10 @@ OsdCpuComputeController::ApplyVertexEdits(
context->GetCurrentVertexBuffer(),
edit->GetPrimvarOffset(),
edit->GetPrimvarWidth(),
batch.vertexOffset,
batch.tableOffset,
batch.start,
batch.end,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
static_cast<unsigned int*>(primvarIndices->GetBuffer()),
static_cast<float*>(editValues->GetBuffer()));
} else if (edit->GetOperation() == FarVertexEdit::Set) {
@ -304,10 +304,10 @@ OsdCpuComputeController::ApplyVertexEdits(
context->GetCurrentVertexBuffer(),
edit->GetPrimvarOffset(),
edit->GetPrimvarWidth(),
batch.vertexOffset,
batch.tableOffset,
batch.start,
batch.end,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
static_cast<unsigned int*>(primvarIndices->GetBuffer()),
static_cast<float*>(editValues->GetBuffer()));
}

34
opensubdiv/osd/cudaComputeController.cpp
View File

@ -134,7 +134,7 @@ OsdCudaComputeController::ApplyBilinearFaceVerticesKernel(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(F_IT->GetCudaMemory()),
static_cast<int*>(F_ITa->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -154,7 +154,7 @@ OsdCudaComputeController::ApplyBilinearEdgeVerticesKernel(
context->GetCurrentVertexNumElements()-3,
context->GetCurrentVaryingNumElements(),
static_cast<int*>(E_IT->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -174,7 +174,7 @@ OsdCudaComputeController::ApplyBilinearVertexVerticesKernel(
context->GetCurrentVertexNumElements()-3,
context->GetCurrentVaryingNumElements(),
static_cast<int*>(V_ITa->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -197,7 +197,7 @@ OsdCudaComputeController::ApplyCatmarkFaceVerticesKernel(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(F_IT->GetCudaMemory()),
static_cast<int*>(F_ITa->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -220,7 +220,7 @@ OsdCudaComputeController::ApplyCatmarkEdgeVerticesKernel(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(E_IT->GetCudaMemory()),
static_cast<float*>(E_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -246,7 +246,7 @@ OsdCudaComputeController::ApplyCatmarkVertexVerticesKernelB(
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<int*>(V_IT->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -269,7 +269,7 @@ OsdCudaComputeController::ApplyCatmarkVertexVerticesKernelA1(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -292,7 +292,7 @@ OsdCudaComputeController::ApplyCatmarkVertexVerticesKernelA2(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -315,7 +315,7 @@ OsdCudaComputeController::ApplyLoopEdgeVerticesKernel(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(E_IT->GetCudaMemory()),
static_cast<float*>(E_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -341,7 +341,7 @@ OsdCudaComputeController::ApplyLoopVertexVerticesKernelB(
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<int*>(V_IT->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -364,7 +364,7 @@ OsdCudaComputeController::ApplyLoopVertexVerticesKernelA1(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -387,7 +387,7 @@ OsdCudaComputeController::ApplyLoopVertexVerticesKernelA2(
context->GetCurrentVaryingNumElements(),
static_cast<int*>(V_ITa->GetCudaMemory()),
static_cast<float*>(V_W->GetCudaMemory()),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -398,7 +398,7 @@ OsdCudaComputeController::ApplyVertexEdits(
static_cast<OsdCudaComputeContext*>(clientdata);
assert(context);
const OsdCudaHEditTable *edit = context->GetEditTable(batch.tableIndex);
const OsdCudaHEditTable *edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
const OsdCudaTable * primvarIndices = edit->GetPrimvarIndices();
@ -410,10 +410,10 @@ OsdCudaComputeController::ApplyVertexEdits(
context->GetCurrentVertexNumElements()-3,
edit->GetPrimvarOffset(),
edit->GetPrimvarWidth(),
batch.vertexOffset,
batch.tableOffset,
batch.start,
batch.end,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
static_cast<int*>(primvarIndices->GetCudaMemory()),
static_cast<float*>(editValues->GetCudaMemory()));
} else if (edit->GetOperation() == FarVertexEdit::Set) {

37
opensubdiv/osd/glslComputeController.cpp
View File

@ -130,7 +130,7 @@ OsdGLSLComputeController::ApplyBilinearFaceVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyBilinearFaceVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -144,7 +144,7 @@ OsdGLSLComputeController::ApplyBilinearEdgeVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyBilinearEdgeVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -158,7 +158,7 @@ OsdGLSLComputeController::ApplyBilinearVertexVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyBilinearVertexVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -172,7 +172,7 @@ OsdGLSLComputeController::ApplyCatmarkFaceVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyCatmarkFaceVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
@ -188,7 +188,7 @@ OsdGLSLComputeController::ApplyCatmarkEdgeVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyCatmarkEdgeVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -202,7 +202,7 @@ OsdGLSLComputeController::ApplyCatmarkVertexVerticesKernelB(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyCatmarkVertexVerticesKernelB(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -216,7 +216,7 @@ OsdGLSLComputeController::ApplyCatmarkVertexVerticesKernelA1(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyCatmarkVertexVerticesKernelA(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -230,7 +230,7 @@ OsdGLSLComputeController::ApplyCatmarkVertexVerticesKernelA2(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyCatmarkVertexVerticesKernelA(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -244,7 +244,7 @@ OsdGLSLComputeController::ApplyLoopEdgeVerticesKernel(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyLoopEdgeVerticesKernel(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -258,7 +258,7 @@ OsdGLSLComputeController::ApplyLoopVertexVerticesKernelB(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyLoopVertexVerticesKernelB(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -272,7 +272,7 @@ OsdGLSLComputeController::ApplyLoopVertexVerticesKernelA1(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyLoopVertexVerticesKernelA(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -286,7 +286,7 @@ OsdGLSLComputeController::ApplyLoopVertexVerticesKernelA2(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
kernelBundle->ApplyLoopVertexVerticesKernelA(
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -299,18 +299,21 @@ OsdGLSLComputeController::ApplyVertexEdits(
OsdGLSLComputeKernelBundle * kernelBundle = context->GetKernelBundle();
const OsdGLSLComputeHEditTable * edit = context->GetEditTable(batch.tableIndex);
const OsdGLSLComputeHEditTable * edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
context->BindEditShaderStorageBuffers(batch.tableIndex);
context->BindEditShaderStorageBuffers(batch.GetTableIndex());
int primvarOffset = edit->GetPrimvarOffset();
int primvarWidth = edit->GetPrimvarWidth();
if (edit->GetOperation() == FarVertexEdit::Add) {
kernelBundle->ApplyEditAdd(primvarOffset, primvarWidth,
batch.vertexOffset, batch.tableOffset,
batch.start, batch.end);
kernelBundle->ApplyEditAdd( primvarOffset,
primvarWidth,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd());
} else {
// XXX: edit SET is not implemented yet.
}

30
opensubdiv/osd/glslTransformFeedbackComputeController.cpp
View File

@ -123,7 +123,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyBilinearFaceVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -141,7 +141,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyBilinearEdgeVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -159,7 +159,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyBilinearVertexVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -177,7 +177,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyCatmarkFaceVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
@ -197,7 +197,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyCatmarkEdgeVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -215,7 +215,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyCatmarkVertexVerticesKernelB(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -233,7 +233,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyCatmarkVertexVerticesKernelA1(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -251,7 +251,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyCatmarkVertexVerticesKernelA2(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -269,7 +269,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyLoopEdgeVerticesKernel(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -287,7 +287,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyLoopVertexVerticesKernelB(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -305,7 +305,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyLoopVertexVerticesKernelA1(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -323,7 +323,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyLoopVertexVerticesKernelA2(
context->GetNumCurrentVertexElements(),
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -336,10 +336,10 @@ OsdGLSLTransformFeedbackComputeController::ApplyVertexEdits(
OsdGLSLTransformFeedbackKernelBundle * kernelBundle = context->GetKernelBundle();
const OsdGLSLTransformFeedbackHEditTable * edit = context->GetEditTable(batch.tableIndex);
const OsdGLSLTransformFeedbackHEditTable * edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
context->BindEditTextures(batch.tableIndex);
context->BindEditTextures(batch.GetTableIndex());
int primvarOffset = edit->GetPrimvarOffset();
int primvarWidth = edit->GetPrimvarWidth();
@ -351,7 +351,7 @@ OsdGLSLTransformFeedbackComputeController::ApplyVertexEdits(
context->GetCurrentVaryingBuffer(),
context->GetNumCurrentVaryingElements(),
primvarOffset, primvarWidth,
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
} else {
// XXX: edit SET is not implemented yet.
}

42
opensubdiv/osd/ompComputeController.cpp
View File

@ -88,7 +88,7 @@ OsdOmpComputeController::ApplyBilinearFaceVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::F_IT)->GetBuffer(),
(const int*)context->GetTable(Table::F_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -104,7 +104,7 @@ OsdOmpComputeController::ApplyBilinearEdgeVerticesKernel(
context->GetCurrentVertexBuffer(),
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -120,7 +120,7 @@ OsdOmpComputeController::ApplyBilinearVertexVerticesKernel(
context->GetCurrentVertexBuffer(),
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -137,7 +137,7 @@ OsdOmpComputeController::ApplyCatmarkFaceVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::F_IT)->GetBuffer(),
(const int*)context->GetTable(Table::F_ITa)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -154,7 +154,7 @@ OsdOmpComputeController::ApplyCatmarkEdgeVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
(const float*)context->GetTable(Table::E_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -172,7 +172,7 @@ OsdOmpComputeController::ApplyCatmarkVertexVerticesKernelB(
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const int*)context->GetTable(Table::V_IT)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -189,7 +189,7 @@ OsdOmpComputeController::ApplyCatmarkVertexVerticesKernelA1(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -206,7 +206,7 @@ OsdOmpComputeController::ApplyCatmarkVertexVerticesKernelA2(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -223,7 +223,7 @@ OsdOmpComputeController::ApplyLoopEdgeVerticesKernel(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::E_IT)->GetBuffer(),
(const float*)context->GetTable(Table::E_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -241,7 +241,7 @@ OsdOmpComputeController::ApplyLoopVertexVerticesKernelB(
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const int*)context->GetTable(Table::V_IT)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd());
}
void
@ -258,7 +258,7 @@ OsdOmpComputeController::ApplyLoopVertexVerticesKernelA1(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, false);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), false);
}
void
@ -275,7 +275,7 @@ OsdOmpComputeController::ApplyLoopVertexVerticesKernelA2(
context->GetCurrentVaryingBuffer(),
(const int*)context->GetTable(Table::V_ITa)->GetBuffer(),
(const float*)context->GetTable(Table::V_W)->GetBuffer(),
batch.vertexOffset, batch.tableOffset, batch.start, batch.end, true);
batch.GetVertexOffset(), batch.GetTableOffset(), batch.GetStart(), batch.GetEnd(), true);
}
void
@ -286,7 +286,7 @@ OsdOmpComputeController::ApplyVertexEdits(
static_cast<OsdCpuComputeContext*>(clientdata);
assert(context);
const OsdCpuHEditTable *edit = context->GetEditTable(batch.tableIndex);
const OsdCpuHEditTable *edit = context->GetEditTable(batch.GetTableIndex());
assert(edit);
const OsdCpuTable * primvarIndices = edit->GetPrimvarIndices();
@ -297,10 +297,10 @@ OsdOmpComputeController::ApplyVertexEdits(
context->GetCurrentVertexBuffer(),
edit->GetPrimvarOffset(),
edit->GetPrimvarWidth(),
batch.vertexOffset,
batch.tableOffset,
batch.start,
batch.end,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
static_cast<unsigned int*>(primvarIndices->GetBuffer()),
static_cast<float*>(editValues->GetBuffer()));
} else if (edit->GetOperation() == FarVertexEdit::Set) {
@ -308,10 +308,10 @@ OsdOmpComputeController::ApplyVertexEdits(
context->GetCurrentVertexBuffer(),
edit->GetPrimvarOffset(),
edit->GetPrimvarWidth(),
batch.vertexOffset,
batch.tableOffset,
batch.start,
batch.end,
batch.GetVertexOffset(),
batch.GetTableOffset(),
batch.GetStart(),
batch.GetEnd(),
static_cast<unsigned int*>(primvarIndices->GetBuffer()),
static_cast<float*>(editValues->GetBuffer()));
}