struct regex *
compile_regex(struct buffer *b, int flags, int needed_sub)
{
struct regex *new_regex;
size_t re_len;
/* // matches the last RE */
if (size_buffer(b) == 0)
{
if (flags > 0)
bad_prog(_(BAD_MODIF));
return NULL;
}
re_len = size_buffer(b);
new_regex = ck_malloc(sizeof (struct regex) + re_len - 1);
new_regex->flags = flags;
memcpy (new_regex->re, get_buffer(b), re_len);
#ifdef REG_PERL
new_regex->sz = re_len;
#else
/* GNU regex does not process \t & co. */
new_regex->sz = normalize_text(new_regex->re, re_len, TEXT_REGEX);
#endif
compile_regex_1 (new_regex, needed_sub);
return new_regex;
}
virtual uint64_t state_buffer_size(device_memory& kg, device_memory& data, size_t num_threads)
{
device_vector<uint64_t> size_buffer(device, "size_buffer", MEM_READ_WRITE);
size_buffer.alloc(1);
size_buffer.zero_to_device();
uint threads = num_threads;
device->kernel_set_args(device->program_state_buffer_size(), 0, kg, data, threads, size_buffer);
size_t global_size = 64;
device->ciErr = clEnqueueNDRangeKernel(device->cqCommandQueue,
device->program_state_buffer_size(),
1,
NULL,
&global_size,
NULL,
0,
NULL,
NULL);
device->opencl_assert_err(device->ciErr, "clEnqueueNDRangeKernel");
size_buffer.copy_from_device(0, 1, 1);
size_t size = size_buffer[0];
size_buffer.free();
if(device->ciErr != CL_SUCCESS) {
string message = string_printf("OpenCL error: %s in clEnqueueNDRangeKernel()",
clewErrorString(device->ciErr));
device->opencl_error(message);
return 0;
}
return size;
}
/* Apply WINDOW to the streams given by APPL. */
static svn_error_t *
apply_window(svn_txdelta_window_t *window, void *baton)
{
struct apply_baton *ab = (struct apply_baton *) baton;
apr_size_t len;
svn_error_t *err;
if (window == NULL)
{
/* We're done; just clean up. */
if (ab->result_digest)
apr_md5_final(ab->result_digest, &(ab->md5_context));
err = svn_stream_close(ab->target);
svn_pool_destroy(ab->pool);
return err;
}
/* Make sure the source view didn't slide backwards. */
SVN_ERR_ASSERT(window->sview_len == 0
|| (window->sview_offset >= ab->sbuf_offset
&& (window->sview_offset + window->sview_len
>= ab->sbuf_offset + ab->sbuf_len)));
/* Make sure there's enough room in the target buffer. */
SVN_ERR(size_buffer(&ab->tbuf, &ab->tbuf_size, window->tview_len, ab->pool));
/* Prepare the source buffer for reading from the input stream. */
if (window->sview_offset != ab->sbuf_offset
|| window->sview_len > ab->sbuf_size)
{
char *old_sbuf = ab->sbuf;
/* Make sure there's enough room. */
SVN_ERR(size_buffer(&ab->sbuf, &ab->sbuf_size, window->sview_len,
ab->pool));
/* If the existing view overlaps with the new view, copy the
* overlap to the beginning of the new buffer. */
if (ab->sbuf_offset + ab->sbuf_len > window->sview_offset)
{
apr_size_t start =
(apr_size_t)(window->sview_offset - ab->sbuf_offset);
memmove(ab->sbuf, old_sbuf + start, ab->sbuf_len - start);
ab->sbuf_len -= start;
}
else
ab->sbuf_len = 0;
ab->sbuf_offset = window->sview_offset;
}
/* Read the remainder of the source view into the buffer. */
if (ab->sbuf_len < window->sview_len)
{
len = window->sview_len - ab->sbuf_len;
err = svn_stream_read(ab->source, ab->sbuf + ab->sbuf_len, &len);
if (err == SVN_NO_ERROR && len != window->sview_len - ab->sbuf_len)
err = svn_error_create(SVN_ERR_INCOMPLETE_DATA, NULL,
"Delta source ended unexpectedly");
if (err != SVN_NO_ERROR)
return err;
ab->sbuf_len = window->sview_len;
}
/* Apply the window instructions to the source view to generate
the target view. */
len = window->tview_len;
svn_txdelta_apply_instructions(window, ab->sbuf, ab->tbuf, &len);
SVN_ERR_ASSERT(len == window->tview_len);
/* Write out the output. */
/* ### We've also considered just adding two (optionally null)
arguments to svn_stream_create(): read_checksum and
write_checksum. Then instead of every caller updating an md5
context when it calls svn_stream_write() or svn_stream_read(),
streams would do it automatically, and verify the checksum in
svn_stream_closed(). But this might be overkill for issue #689;
so for now we just update the context here. */
if (ab->result_digest)
apr_md5_update(&(ab->md5_context), ab->tbuf, len);
return svn_stream_write(ab->target, ab->tbuf, &len);
}
/* Apply WINDOW to the streams given by APPL. */
static svn_error_t *
apply_window(svn_txdelta_window_t *window, void *baton)
{
struct apply_baton *ab = (struct apply_baton *) baton;
apr_size_t len;
svn_error_t *err;
if (window == NULL)
{
/* We're done; just clean up. */
if (ab->result_digest)
apr_md5_final(ab->result_digest, &(ab->md5_context));
err = svn_stream_close(ab->target);
svn_pool_destroy(ab->pool);
return err;
}
/* Make sure the source view didn't slide backwards. */
SVN_ERR_ASSERT(window->sview_len == 0
|| (window->sview_offset >= ab->sbuf_offset
&& (window->sview_offset + window->sview_len
>= ab->sbuf_offset + ab->sbuf_len)));
/* Make sure there's enough room in the target buffer. */
SVN_ERR(size_buffer(&ab->tbuf, &ab->tbuf_size, window->tview_len, ab->pool));
/* Prepare the source buffer for reading from the input stream. */
if (window->sview_offset != ab->sbuf_offset
|| window->sview_len > ab->sbuf_size)
{
char *old_sbuf = ab->sbuf;
/* Make sure there's enough room. */
SVN_ERR(size_buffer(&ab->sbuf, &ab->sbuf_size, window->sview_len,
ab->pool));
/* If the existing view overlaps with the new view, copy the
* overlap to the beginning of the new buffer. */
if ( (apr_size_t)ab->sbuf_offset + ab->sbuf_len
> (apr_size_t)window->sview_offset)
{
apr_size_t start =
(apr_size_t)(window->sview_offset - ab->sbuf_offset);
memmove(ab->sbuf, old_sbuf + start, ab->sbuf_len - start);
ab->sbuf_len -= start;
}
else
ab->sbuf_len = 0;
ab->sbuf_offset = window->sview_offset;
}
/* Read the remainder of the source view into the buffer. */
if (ab->sbuf_len < window->sview_len)
{
len = window->sview_len - ab->sbuf_len;
err = svn_stream_read_full(ab->source, ab->sbuf + ab->sbuf_len, &len);
if (err == SVN_NO_ERROR && len != window->sview_len - ab->sbuf_len)
err = svn_error_create(SVN_ERR_INCOMPLETE_DATA, NULL,
"Delta source ended unexpectedly");
if (err != SVN_NO_ERROR)
return err;
ab->sbuf_len = window->sview_len;
}
/* Apply the window instructions to the source view to generate
the target view. */
len = window->tview_len;
svn_txdelta_apply_instructions(window, ab->sbuf, ab->tbuf, &len);
SVN_ERR_ASSERT(len == window->tview_len);
/* Write out the output. */
/* Just update the context here. */
if (ab->result_digest)
apr_md5_update(&(ab->md5_context), ab->tbuf, len);
return svn_stream_write(ab->target, ab->tbuf, &len);
}