commandqueue.cpp 9.32 KB
Newer Older
schultezub's avatar
schultezub committed
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
// ================================================================================================
// 
// This file is part of the KissCL, an OpenCL C++ wrapper following the KISS principle.
// 
// Copyright (C) 2012, all rights reserved,
//      Christian Schulte zu Berge (software@cszb.net)
// 
// This library is free software; you can redistribute it and/or modify it under the terms of the 
// GNU Lesser General Public License version 3 as published by the Free Software Foundation.
// 
// This library is distributed in the hope that it will be useful, but WITHOUT ANY WARRANTY; 
// without even the implied warranty of MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See 
// the GNU Lesser General Public License for more details.
// 
// You should have received a copy of the GNU Lesser General Public License in the file 
// "LICENSE.txt" along with this library. If not, see <http://www.gnu.org/licenses/>.
// 
// ================================================================================================

20
21
22
23
#include "commandqueue.h"

#include "tgt/assert.h"
#include "tgt/logmanager.h"
schultezub's avatar
schultezub committed
24
25
26
27
28
29
#include "kisscl/context.h"
#include "kisscl/device.h"
#include "kisscl/event.h"
#include "kisscl/kernel.h"
#include "kisscl/memory.h"
#include "kisscl/platform.h"
30

schultezub's avatar
schultezub committed
31
namespace kisscl {
32

schultezub's avatar
schultezub committed
33
    const std::string CommandQueue::loggerCat_ = "kisscl.CommandQueue";
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65

    CommandQueue::CommandQueue(Context* context, cl_command_queue_properties properties /*= 0*/)
        : CLWrapper<cl_command_queue>(0)
        , _context(context)
        , _device(context->getDevices().front())
        , _profilingEnabled((properties & CL_QUEUE_PROFILING_ENABLE) != 0)
    {
        tgtAssert(_context != 0, "Context must not be 0.");
        tgtAssert(_device != 0, "Device must not be 0. Something went terribly wrong, this should should have been asserted earlier.");

        cl_int err;
        _id = clCreateCommandQueue(_context->getId(), _device->getId(), properties, &err);
        LCL_ERROR(err);
    }

    CommandQueue::CommandQueue(Context* context, Device* device, cl_command_queue_properties properties /*= 0*/)
        : CLWrapper<cl_command_queue>(0)
        , _context(context)
        , _device(device)
        , _profilingEnabled((properties & CL_QUEUE_PROFILING_ENABLE) != 0)
    {
        tgtAssert(_context != 0, "Context must not be 0.");
        tgtAssert(_device != 0, "Device must not be 0.");

        cl_int err;
        _id = clCreateCommandQueue(_context->getId(), _device->getId(), properties, &err);
        LCL_ERROR(err);
    }

// = getters and setters ==========================================================================

    CommandQueue::~CommandQueue() {
schultezub's avatar
schultezub committed
66

67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
    }

    const Context* CommandQueue::getContext() const {
        return _context;
    }

    const Device* CommandQueue::getDevice() const {
        return _device;
    }

// = the interesting stuff :) =====================================================================

    void CommandQueue::flush() {
        LCL_ERROR(clFlush(_id));
    }

    void CommandQueue::finish() {
        LCL_ERROR(clFinish(_id));
    }

87
    Event CommandQueue::enqueueTask(const Kernel* kernel, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
88
89
90
        tgtAssert(kernel != 0, "Kernel must not be 0.");

        cl_event e;
91
        LCL_ERROR(clEnqueueTask(_id, kernel->getId(), eventsToWaitFor._size, eventsToWaitFor._items, &e));
92
93
94
        return Event(e);
    }

95
    Event CommandQueue::enqueueKernel(const Kernel* kernel, size_t globalWorkSize, size_t localWorkSize /*= 0*/, size_t offset /*= 0*/, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
96
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
97
98
        tgtAssert(localWorkSize == 0 || localWorkSize < globalWorkSize, "Global work size must be greater than local work size.");
        tgtAssert(localWorkSize == 0 || (globalWorkSize % localWorkSize == 0), "Global work size must be a multiple than local work size.");
99
100
101
102
103
104
105
106
107
108

        cl_event e;
        LCL_ERROR(clEnqueueNDRangeKernel(
            _id, 
            kernel->getId(), 
            1, 
            (offset == 0 ? 0 : &offset), 
            &globalWorkSize, 
            (localWorkSize == 0 ? 0 : &localWorkSize), 
            eventsToWaitFor._size, 
109
            eventsToWaitFor._items, 
110
111
112
113
            &e));
        return Event(e);
    }

114
    Event CommandQueue::enqueueKernel(const Kernel* kernel, tgt::svec2 globalWorkSize, tgt::svec2 localWorkSize /*= tgt::svec2::zero*/, tgt::svec2 offset /*= tgt::svec2::zero*/, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
115
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
116
117
118
        tgtAssert(localWorkSize == tgt::svec2::zero || tgt::hor(tgt::lessThan(localWorkSize, globalWorkSize)), "Global work size must be greater than local work size.");
        tgtAssert(localWorkSize == tgt::svec2::zero || (globalWorkSize.x % localWorkSize.x == 0), "Global work size must be a multiple than local work size.");
        tgtAssert(localWorkSize == tgt::svec2::zero || (globalWorkSize.y % localWorkSize.y == 0), "Global work size must be a multiple than local work size.");
119
120
121
122
123
124
125
126
127
128

        cl_event e;
        LCL_ERROR(clEnqueueNDRangeKernel(
            _id, 
            kernel->getId(), 
            2, 
            (offset == tgt::svec2::zero ? 0 : offset.elem), 
            globalWorkSize.elem,
            (localWorkSize == tgt::svec2::zero ? 0 : localWorkSize.elem),
            eventsToWaitFor._size, 
129
            eventsToWaitFor._items, 
130
131
132
133
            &e));
        return Event(e);
    }

134
    Event CommandQueue::enqueueKernel(const Kernel* kernel, tgt::svec3 globalWorkSize, tgt::svec3 localWorkSize /*= tgt::svec3::zero*/, tgt::svec3 offset /*= tgt::svec3::zero*/, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
135
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
136
137
138
139
        tgtAssert(localWorkSize == tgt::svec3::zero || tgt::hor(tgt::lessThan (localWorkSize, globalWorkSize)), "Global work size must be greater than local work size.");
        tgtAssert(localWorkSize == tgt::svec3::zero || (globalWorkSize.x % localWorkSize.x == 0), "Global work size must be a multiple than local work size.");
        tgtAssert(localWorkSize == tgt::svec3::zero || (globalWorkSize.y % localWorkSize.y == 0), "Global work size must be a multiple than local work size.");
        tgtAssert(localWorkSize == tgt::svec3::zero || (globalWorkSize.z % localWorkSize.z == 0), "Global work size must be a multiple than local work size.");
140
141
142
143
144
145
146
147
148
149

        cl_event e;
        LCL_ERROR(clEnqueueNDRangeKernel(
            _id, 
            kernel->getId(), 
            3, 
            (offset == tgt::svec3::zero ? 0 : offset.elem), 
            globalWorkSize.elem,
            (localWorkSize == tgt::svec3::zero ? 0 : localWorkSize.elem),
            eventsToWaitFor._size, 
150
            eventsToWaitFor._items, 
151
152
153
154
            &e));
        return Event(e);
    }

155
156
157
158
159
160
161
162
163
164
    void CommandQueue::enqueueBarrier() {
        LCL_ERROR(clEnqueueBarrier(_id));
    }

    Event CommandQueue::enqueueMarker() {
        cl_event e;
        LCL_ERROR(clEnqueueMarker(_id, &e));
        return Event(e);
    }

165
166
    void CommandQueue::enqueueWaitForEvents(const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
        LCL_ERROR(clEnqueueWaitForEvents(_id, eventsToWaitFor._size, eventsToWaitFor._items));
167
168
    }

169
    Event CommandQueue::enqueueRead(const Buffer* buffer, void* data, bool blocking /*= true*/, size_t offset /*= 0*/, size_t numBytes /*= 0*/, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
170
171
        cl_event e;
        if (numBytes == 0) {
172
            LCL_ERROR(clEnqueueReadBuffer(_id, buffer->getId(), blocking, offset, buffer->getSize(), data, eventsToWaitFor._size, eventsToWaitFor._items, &e));
173
174
        }
        else {
175
            LCL_ERROR(clEnqueueReadBuffer(_id, buffer->getId(), blocking, offset, numBytes, data, eventsToWaitFor._size, eventsToWaitFor._items, &e));
176
177
178
179
        }
        return Event(e);
    }

180
    Event CommandQueue::enqueueWrite(const Buffer* buffer, void* data, bool blocking /*= true*/, size_t offset /*= 0*/, size_t numBytes /*= 0*/, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
181
182
        cl_event e;
        if (numBytes == 0) {
183
            LCL_ERROR(clEnqueueWriteBuffer(_id, buffer->getId(), blocking, offset, buffer->getSize(), data, eventsToWaitFor._size, eventsToWaitFor._items, &e));
184
185
        }
        else {
186
            LCL_ERROR(clEnqueueWriteBuffer(_id, buffer->getId(), blocking, offset, numBytes, data, eventsToWaitFor._size, eventsToWaitFor._items, &e));
187
188
189
190
        }
        return Event(e);
    }

191
    Event CommandQueue::enqueueAcquireGLObject(const ItemList<GLTexture>& textures, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
schultezub's avatar
schultezub committed
192
        cl_event e;
193
        LCL_ERROR(clEnqueueAcquireGLObjects(_id, textures._size, textures._items, eventsToWaitFor._size, eventsToWaitFor._items, &e));
schultezub's avatar
schultezub committed
194
195
196
        return Event(e);
    }

197
    Event CommandQueue::enqueueReleaseGLObject(const ItemList<GLTexture>& textures, const ItemList<Event>& eventsToWaitFor /*= ItemList<Event>()*/) {
schultezub's avatar
schultezub committed
198
        cl_event e;
199
        LCL_ERROR(clEnqueueReleaseGLObjects(_id, textures._size, textures._items, eventsToWaitFor._size, eventsToWaitFor._items, &e));
schultezub's avatar
schultezub committed
200
201
202
        return Event(e);
    }

203

204
205

}