commandqueue.cpp 9.31 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
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97

    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() {
        if (_id != 0)
            clReleaseCommandQueue(_id);
    }

    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));
    }

    Event CommandQueue::enqueueTask(const Kernel* kernel, const EventList& eventsToWaitFor /*= EventList()*/) {
        tgtAssert(kernel != 0, "Kernel must not be 0.");

        cl_event e;
        LCL_ERROR(clEnqueueTask(_id, kernel->getId(), eventsToWaitFor._size, eventsToWaitFor._events, &e));
        return Event(e);
    }

    Event CommandQueue::enqueueKernel(const Kernel* kernel, size_t globalWorkSize, size_t localWorkSize /*= 0*/, size_t offset /*= 0*/, const EventList& eventsToWaitFor /*= EventList()*/) {
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
98
99
        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.");
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116

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

    Event CommandQueue::enqueueKernel(const Kernel* kernel, tgt::svec2 globalWorkSize, tgt::svec2 localWorkSize /*= tgt::svec2::zero*/, tgt::svec2 offset /*= tgt::svec2::zero*/, const EventList& eventsToWaitFor /*= EventList()*/) {
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
117
118
119
        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.");
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136

        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, 
            eventsToWaitFor._events, 
            &e));
        return Event(e);
    }

    Event CommandQueue::enqueueKernel(const Kernel* kernel, tgt::svec3 globalWorkSize, tgt::svec3 localWorkSize /*= tgt::svec3::zero*/, tgt::svec3 offset /*= tgt::svec3::zero*/, const EventList& eventsToWaitFor /*= EventList()*/) {
        tgtAssert(kernel != 0, "Kernel must not be 0.");
schultezub's avatar
schultezub committed
137
138
139
140
        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.");
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155

        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, 
            eventsToWaitFor._events, 
            &e));
        return Event(e);
    }

156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
182
183
184
185
186
187
188
189
190
191
    void CommandQueue::enqueueBarrier() {
        LCL_ERROR(clEnqueueBarrier(_id));
    }

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

    void CommandQueue::enqueueWaitForEvents(const EventList& eventsToWaitFor /*= EventList()*/) {
        LCL_ERROR(clEnqueueWaitForEvents(_id, eventsToWaitFor._size, eventsToWaitFor._events));
    }

    Event CommandQueue::enqueueRead(const Buffer* buffer, void* data, bool blocking /*= true*/, size_t offset /*= 0*/, size_t numBytes /*= 0*/, const EventList& eventsToWaitFor /*= EventList()*/) {
        cl_event e;
        if (numBytes == 0) {
            LCL_ERROR(clEnqueueReadBuffer(_id, buffer->getId(), blocking, offset, buffer->getSize(), data, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        }
        else {
            LCL_ERROR(clEnqueueReadBuffer(_id, buffer->getId(), blocking, offset, numBytes, data, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        }
        return Event(e);
    }

    Event CommandQueue::enqueueWrite(const Buffer* buffer, void* data, bool blocking /*= true*/, size_t offset /*= 0*/, size_t numBytes /*= 0*/, const EventList& eventsToWaitFor /*= EventList()*/) {
        cl_event e;
        if (numBytes == 0) {
            LCL_ERROR(clEnqueueWriteBuffer(_id, buffer->getId(), blocking, offset, buffer->getSize(), data, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        }
        else {
            LCL_ERROR(clEnqueueWriteBuffer(_id, buffer->getId(), blocking, offset, numBytes, data, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        }
        return Event(e);
    }

schultezub's avatar
schultezub committed
192
193
194
195
196
197
198
199
200
201
202
203
204
205
    Event CommandQueue::enqueueAcquireGLObject(const SharedTexture* texture, const EventList& eventsToWaitFor /*= EventList()*/) {
        cl_event e;
        cl_mem mem = texture->getId();
        LCL_ERROR(clEnqueueAcquireGLObjects(_id, 1, &mem, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        return Event(e);
    }

    Event CommandQueue::enqueueReleaseGLObject(const SharedTexture* texture, const EventList& eventsToWaitFor /*= EventList()*/) {
        cl_event e;
        cl_mem mem = texture->getId();
        LCL_ERROR(clEnqueueReleaseGLObjects(_id, 1, &mem, eventsToWaitFor._size, eventsToWaitFor._events, &e));
        return Event(e);
    }

206

207
208

}