GPU, Python and Maya

team470

Here an example how to use pyopencl in Maya.
yTwistNodeGPU.py
# --------------------------------------------------------------------------------
# Copyright (c) 2013 Mack Stone. All rights reserved.
#
# Permission is hereby granted, free of charge, to any person obtaining a copy
# of this software and associated documentation files (the "Software"), to deal
# in the Software without restriction, including without limitation the rights
# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
# copies of the Software, and to permit persons to whom the Software is
# furnished to do so, subject to the following conditions:
#
# The above copyright notice and this permission notice shall be included in
# all copies or substantial portions of the Software.
#
# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
# THE SOFTWARE.
# --------------------------------------------------------------------------------
"""
Simple deform node use pyopencl in Maya.
@author: Mack Stone
"""
import sys
import logging
import maya.OpenMaya as om
import maya.OpenMayaMPx as ompx
from maya import utils
import numpy
import pyopencl as cl

# opencl code, run on GPU
kernelCode = """
__kernel void ytwist(__global const float4 *pos,
__global float4 *newPos,
float angle,
float envelope)
{
int gid = get_global_id(0);
newPos[gid].xyzw = pos[gid].xyzw;
float ff = angle * pos[gid].y * envelope;
if(ff != 0.0f)
{
float cct = cos(ff);
float cst = sin(ff);
newPos[gid].x = pos[gid].x * cct - pos[gid].z * cst;
newPos[gid].z = pos[gid].x * cst + pos[gid].z * cct;
}
}
"""
class YTwistNode(ompx.MPxDeformerNode):
NAME = "yTwistNode"
ID = om.MTypeId(0x8702)
angle = om.MObject()
def __init__(self):
ompx.MPxDeformerNode.__init__(self)
# create context
self._ctx = cl.create_some_context()
# command queue
self._queue = cl.CommandQueue(self._ctx)
# create and build GPU program
self._program = cl.Program(self._ctx, kernelCode).build()
# setup logger
formatter = logging.Formatter("%(asctime)s - %(message)s")
utils._guiLogHandler.setFormatter(formatter)
def deform(self, dataBlock, geomIter, matrix, multiIndex):
logging.info("start deforming")
# get the angle from the datablock
angleHandle = dataBlock.inputValue( self.angle )
angleValue = angleHandle.asDouble()
# get the envelope
envelope = OpenMayaMPx.cvar.MPxDeformerNode_envelope
envelopeHandle = dataBlock.inputValue( envelope )
envelopeValue = envelopeHandle.asFloat()
# get all position data
logging.info("get all position data")
pos = numpy.zeros((geomIter.count(), 4), dtype=numpy.float32)
while not geomIter.isDone():
point = geomIter.position()
index = geomIter.index()
pos[index, 0] = point.x
pos[index, 1] = point.y
pos[index, 2] = point.z
pos[index, 3] = point.w
geomIter.next()
logging.info("start copy data to GPU")
memf = cl.mem_flags
# create buffer from pos
posBuf = cl.Buffer(self._ctx, memf.READ_ONLY | memf.COPY_HOST_PTR, hostbuf=pos)
# create write buffer
outBuf = cl.Buffer(self._ctx, memf.WRITE_ONLY, pos.nbytes)
# run GPU Program
logging.info("run GPU Program")
self._program.ytwist(self._queue, pos.shape, None, posBuf, outBuf,
numpy.float32(angleValue), numpy.float32(envelopeValue))
logging.info("end GPU Program")
# copy data back to memory
newPos = numpy.zeros_like(pos)
cl.enqueue_copy(self._queue, newPos, outBuf).wait()
# set positions
logging.info("set all position")
geomIter.reset()
while not geomIter.isDone():
point = geomIter.position()
index = geomIter.index()
point.x = float(newPos[index, 0])
point.y = float(newPos[index, 1])
point.z = float(newPos[index, 2])
geomIter.next()
logging.info("end deform")
@staticmethod
def creator():
return ompx.asMPxPtr(YTwistNode())
@staticmethod
def initialize():
# angle
nAttr = om.MFnNumericAttribute()
YTwistNode.angle = nAttr.create("angle", "fa", om.MFnNumericData.kDouble, 0.0)
nAttr.setKeyable(1)
# add attribute
YTwistNode.addAttribute(YTwistNode.angle)
outputGeom = ompx.cvar.MPxDeformerNode_outputGeom
YTwistNode.attributeAffects(YTwistNode.angle, outputGeom)
# initialize the script plug-in
def initializePlugin(mobject):
mplugin = ompx.MFnPlugin(mobject)
try:
mplugin.registerNode(YTwistNode.NAME, YTwistNode.ID, YTwistNode.creator,
YTwistNode.initialize, ompx.MPxNode.kDeformerNode )
except:
sys.stderr.write("Failed to register node: %s\n" % YTwistNode.NAME)
# uninitialize the script plug-in
def uninitializePlugin(mobject):
mplugin = ompx.MFnPlugin(mobject)
try:
mplugin.deregisterNode(YTwistNode.ID)
except:
sys.stderr.write("Failed to unregister node: %s\n" % YTwistNode.NAME )


It's slower than yTwistNode.py(CPU) when I test it.
My testing hardware:
CPU: intel Xeon w3530 2.80GHz
Graphics card: nvidia quadro FX 1800
Test on a obj with 960002 vertexes.
GPU: about 18s a time
CPU: about 10s a time
But the slow part is not the GPU computing. It's the while loop. Two while loop take almost 18s.